跳转至

clEnqueueNDRangeKernel

clEnqueueNDRangeKernel 是 OpenCL 中最常用的函数之一，用于将内核调度到设备上执行。它的作用是定义工作项的分布（全局工作区和局部工作组），并将内核任务加入到命令队列中。

以下是该函数的完整参数说明以及用法。

函数原型

C
cl_int clEnqueueNDRangeKernel(
    cl_command_queue command_queue,  // 命令队列
    cl_kernel kernel,                // 要执行的内核对象
    cl_uint work_dim,                // 工作区的维度 (1D, 2D, 3D)
    const size_t *global_work_offset,// 全局偏移量 (通常为 NULL)
    const size_t *global_work_size,  // 全局工作区大小
    const size_t *local_work_size,   // 局部工作区大小 (工作组大小)
    cl_uint num_events_in_wait_list, // 等待的事件数
    const cl_event *event_wait_list, // 等待的事件列表
    cl_event *event                  // 返回的事件对象
);

参数说明

1. command_queue

  • 类型：cl_command_queue
  • 描述：指定命令队列，表示在哪个设备和上下文中执行内核。
  • 命令队列是通过 clCreateCommandQueue 创建的，绑定到某个设备。

2. kernel

  • 类型：cl_kernel
  • 描述：指定要执行的内核对象。
  • 内核对象是通过 clCreateKernel 从编译好的程序对象中创建的。
  • 在调用此函数之前，必须通过 clSetKernelArg 设置内核的参数。

3. work_dim

  • 类型：cl_uint
  • 描述：指定工作区的维度，可以是 1、2 或 3，表示 1D、2D 或 3D 的工作区。
  • 例如：
  • work_dim = 1：表示 1D 工作区。
  • work_dim = 2：表示 2D 工作区。
  • work_dim = 3：表示 3D 工作区。

4. global_work_offset

  • 类型：const size_t *
  • 描述：全局工作区的偏移量数组，指定全局 ID 的起始偏移值。
  • 通常设置为 NULL，表示从 (0, 0, 0) 开始。
  • 如果设置非 NULL，会将全局 ID 偏移。例如，如果偏移为 (2, 2)，则 get_global_id(0) 的值从 2 开始。

5. global_work_size

  • 类型：const size_t *
  • 描述：全局工作区的大小数组，指定每个维度上工作项的总数。
  • 必须是一个长度为 work_dim 的数组。
  • 例如：
  • 如果 work_dim = 1，global_work_size[0] = 1024，表示 1D 工作区有 1024 个工作项。
  • 如果 work_dim = 2，global_work_size = {16, 16}，表示 2D 工作区大小为 16x16，总共 256 个工作项。
  • 注意：全局工作区大小通常是局部工作区大小的整数倍。

6. local_work_size

  • 类型：const size_t *
  • 描述：局部工作区（工作组）的大小数组，指定每个工作组中工作项的数量。
  • 例如：
  • 如果 work_dim = 1，local_work_size[0] = 64，表示每个工作组包含 64 个工作项。
  • 如果 work_dim = 2，local_work_size = {4, 4}，表示每个工作组大小为 4x4。
  • 注意：
  • 如果设置为 NULL，OpenCL 会自动决定局部工作组大小。
  • 如果指定非 NULL 值，则 global_work_size 必须是 local_work_size 的整数倍，否则会报错。

7. num_events_in_wait_list

  • 类型：cl_uint
  • 描述：指定事件等待列表中的事件数量。
  • 如果设置为 0，表示不需要等待任何事件。

8. event_wait_list

  • 类型：const cl_event *
  • 描述：事件等待列表，表示本次内核执行需要等待的事件。
  • 如果 num_events_in_wait_list = 0，此参数必须为 NULL。
  • 如果指定非 NULL 值，则内核执行会在等待列表中的事件完成后开始。

9. event

  • 类型：cl_event *
  • 描述：用于返回与内核执行相关联的事件对象。
  • 如果不需要事件，可设置为 NULL。
  • 返回的事件可用于查询内核执行状态或作为其他命令的依赖条件。

返回值

  • 类型：cl_int
  • 描述：返回错误代码，常见返回值如下：
  • CL_SUCCESS：函数调用成功。
  • CL_INVALID_COMMAND_QUEUE：无效的命令队列。
  • CL_INVALID_KERNEL：内核对象无效。
  • CL_INVALID_CONTEXT：命令队列和内核对象不属于同一上下文。
  • CL_INVALID_WORK_DIMENSION：工作区维度无效（超出设备支持的范围）。
  • CL_INVALID_GLOBAL_WORK_SIZE：全局工作区大小无效。
  • CL_INVALID_WORK_GROUP_SIZE：局部工作区大小不符合设备限制。
  • CL_OUT_OF_RESOURCES：设备资源不足，无法执行命令。
  • CL_OUT_OF_HOST_MEMORY：主机内存不足。

工作区参数之间的关系

  • 全局工作区大小 (global_work_size)： 总的工作项数量，决定计算任务的规模。

  • 局部工作区大小 (local_work_size)： 每个工作组中的工作项数量，决定工作组的划分方式。

  • 工作组数量： 工作组数量可以通过以下公式计算：

    Text Only
    num_work_groups = global_work_size / local_work_size
    例如：

  • 如果 global_work_size = 1024，local_work_size = 64，则工作组数量为 1024 / 64 = 16。

示例代码

1. 一维工作区示例

C
size_t global_work_size = 1024; // 全局工作区大小
size_t local_work_size = 64;   // 每个工作组大小

cl_int err = clEnqueueNDRangeKernel(
    command_queue, kernel, 1, NULL,
    &global_work_size, &local_work_size,
    0, NULL, NULL
);

if (err != CL_SUCCESS) {
    printf("Error: %d\n", err);
}

2. 二维工作区示例

C
size_t global_work_size[2] = {16, 16}; // 全局工作区大小
size_t local_work_size[2] = {4, 4};   // 局部工作组大小

cl_int err = clEnqueueNDRangeKernel(
    command_queue, kernel, 2, NULL,
    global_work_size, local_work_size,
    0, NULL, NULL
);

if (err != CL_SUCCESS) {
    printf("Error: %d\n", err);
}

3. 自动选择局部工作组大小

C
size_t global_work_size[2] = {16, 16}; // 全局工作区大小

cl_int err = clEnqueueNDRangeKernel(
    command_queue, kernel, 2, NULL,
    global_work_size, NULL,
    0, NULL, NULL
);

if (err != CL_SUCCESS) {
    printf("Error: %d\n", err);
}

注意事项

  1. 设备限制

  2. 局部工作组大小不能超过设备的最大工作组大小，可以通过 clGetDeviceInfo 查询：

    C
    clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE, ...);

  3. 性能优化

  4. 局部工作组大小应尽量与设备计算单元对齐（如 GPU 的线程块）。

  5. 常用的局部工作组大小是 32、64、128 等。具体选择需要根据设备架构优化。

  6. 边界处理

  7. 如果 global_work_size 不是 local_work_size 的整数倍，可能需要在内核代码中处理越界问题。

clEnqueueNDRangeKernel 是 OpenCL 中的关键函数，设置正确的工作区参数对于性能优化和代码正确性至关重要。