Description
The XCL_LATENCY attribute specifies a minimum, or maximum latency value, or both, for the completion of functions, loops, and regions. Latency is defined as the number of clock cycles required to produce an output. Function or region latency is the number of clock cycles required for the code to compute all output values, and return. Loop latency is the number of cycles to execute all iterations of the loop. See Performance Metrics Example of Vitis High-Level Synthesis User Guide (UG1399).
The Vitis HLS tool always tries to minimize latency in the design. When the XCL_LATENCY attribute is specified, the tool behavior is as follows:
- When latency is greater than the minimum, or less than the maximum: The constraint is satisfied. No further optimizations are performed.
- When latency is less than the minimum: If the HLS tool can achieve less than the minimum specified latency, it extends the latency to the specified value, potentially increasing sharing.
- When latency is greater than the maximum: If the HLS tool cannot schedule within the maximum limit, it increases effort to achieve the specified constraint. If it still fails to meet the maximum latency, it issues a warning, and produces a design with the smallest achievable latency in excess of the maximum.
Syntax
__attribute__((xcl_latency(min, max)))Where:
- <min>: Specifies the minimum latency for the function, loop, or region of code.
- <max>: Specifies the maximum latency for the function, loop, or region of code.
Examples
The for loop in the test function is specified to have a minimum latency of 4 and a
maximum latency of 8.
__kernel void test(__global float *A, __global float *B, __global float *C, int id)
{
for (unsigned int i = 0; i < id; i++)
__attribute__((xcl_latency(4, 12))) {
C[id] = A[id] * B[id];
}
}