OpenCL 2.1 NEO设备编译enqueue_kernel报标识符未声明错误

问题环境 #

• 运行设备：Intel(R) Gen9 HD Graphics NEO -- OpenCL 2.1 NEO
• 问题表现：编译内核时，ndrange_t、enqueue_kernel、get_default_queue等设备端入队（动态并行）相关的类型、函数全部提示未声明，同时块语法位置报语法错误

问题内核代码 #

__kernel void update(
    const __global uint* positions,
    const __global float3* offsets, 
    const int size,
    __global int* cost

) {

    int global_id = get_global_id(0);

    if (global_id >= size) {
        return;
    }

    int3 update_index = position_from_index(grid_centroid_positions[global_id], SIZE) -
                        offset_grid;
    
    ndrange_t ndrange = ndrange_3d(size, size, size);

    enqueue_kernel(get_default_queue(), ndrange,
                           ^{update_surrouding_cells(offsets, global_id, update_index, update_edge_size, size, cost)});
}

编译报错信息 #

6:158:5: error: use of undeclared identifier 'ndrange_t'
    ndrange_t ndrange = ndrange_3d(size, size, size);
    ^
6:161:5: error: implicit declaration of function 'enqueue_kernel' is invalid in OpenCL
    enqueue_kernel(get_default_queue(), ndrange,
    ^
6:161:20: error: implicit declaration of function 'get_default_queue' is invalid in OpenCL
    enqueue_kernel(get_default_queue(), ndrange,
                   ^
6:162:163: error: expected ';' after expression
                           ^{update_surrouding_cells(offsets, global_id, update_index, update_edge_size, size, cost)});
                                                                                                                                                                  ^
                                                                                                                                                                  ;
6:161:41: error: use of undeclared identifier 'ndrange'
    enqueue_kernel(get_default_queue(), ndrange,

当前使用的编译选项 #

-I "/home/development/cl" -g
 -D SIZE=256

问题原因与解决方案 #

这个问题不需要启用任何额外扩展，核心原因是编译时未指定OpenCL C语言标准为2.1版本。
包括Intel NEO在内的绝大多数OpenCL平台编译器，默认不会自动使用硬件支持的最高语言标准编译内核，通常会默认fallback到OpenCL C 1.2标准。而设备侧动态入队接口、ndrange_t类型、^{}形式的块（block）语法都是OpenCL C 2.1标准引入的核心特性，在1.2标准下自然会被识别为未声明标识符、非法语法。

修复方法 #

在编译选项中添加-cl-std=CL2.1参数，强制编译器按照OpenCL C 2.1标准解析内核代码即可。
修改后完整的编译选项如下：

-I "/home/development/cl" -g -D SIZE=256 -cl-std=CL2.1

补充说明 #

• 之前报的expected ';' after expression语法错误，本质是编译器不识别OpenCL C 2.1的块语法导致的，指定正确编译标准后该错误会同步解决
• 上述设备侧入队相关特性属于OpenCL 2.1核心规范内容，不需要额外添加#pragma OPENCL EXTENSION相关声明启用扩展
• 如果后续需要使用更高版本OpenCL标准的特性（比如OpenCL C 3.0），同理需要添加对应的-cl-std=CL3.0参数，否则编译器不会默认启用高版本标准的语法和接口

内容的提问来源于stack exchange，提问作者Fantastic Mr Fox