Intel® FPGA SDK for OpenCL™ Standard Edition: Programming Guide

ID 683342
Date 4/22/2019
Public
Document Table of Contents

12.3.1.1. Omit Hardware that Generates and Dispatches Kernel IDs

The max_global_work_dim(0) kernel attribute instructs the Intel® FPGA SDK for OpenCL™ Offline Compiler to omit logic that generates and dispatches global, local, and group IDs into the compiled kernel.

Semantically, the max_global_work_dim(0) kernel attribute specifies that the global work dimension of the kernel is zero. Setting this kernel attribute means that the kernel does not use any global, local, or group IDs. The presence of this attribute in the kernel code serves as a guarantee to the offline compiler that the kernel is a single work-item kernel.

When compiling the following kernel, the offline compiler will generate interface hardware as illustrated in Intel® FPGA SDK for OpenCL™ Offline Compiler-Generated Interface Hardware for a Kernel with the max_global_work_dim(0) Attribute.

channel int chan_in;
channel int chan_out;

__attribute__((max_global_work_dim(0)))
__kernel void plusK (int N, int k) {
    for (int i = 0; i < N; ++i) {
	    int data_in = read_channel_intel(chan_in);
	    write_channel_intel(chan_out, data_in + k);	
    }
}
Figure 19.  Intel® FPGA SDK for OpenCL™ Offline Compiler-Generated Interface Hardware for a Kernel with the max_global_work_dim(0) Attribute

If your current kernel implementation has multiple work-items but does not use global, local, or group IDs, you can use the max_global_work_dim(0) kernel attribute if you modify the kernel code accordingly:

  1. Wrap the kernel body in a for loop that iterates as many times as the number of work-items.
  2. Launch the modified kernel with only one work-item.