如何正确初始化此 C++ for openCL 内核的输入/输出参数？

Question

这是我第一次编写 OpenCL 计算单元，所以我从小处着手；这是我的基本测试内核：

kernel void test_kernel(global float* in, global float* out)
{
    int thread_id = get_global_id(0);
    printf("%d", thread_id);
    out[thread_id] = in[thread_id] + thread_id;
}

下面是尝试为参数构造缓冲区并运行它的 c++ 代码：

...
...

cl::Kernel kernel(program, "test_kernel", &cl_error);
if (cl_error != 0) {
    std::cout << "Error - cl::Kernel - " << getErrorString(cl_error) << std::endl;
    return 1;
}

cl::CommandQueue command_queue(context, device);

cl::vector<float> input_vector{ 0.1f, 0.2f, 0.3f, 0.4f, 0.5f };
cl::vector<float> output_vector{ 0.0f, 0.0f, 0.0f, 0.0f, 0.0f };
cl::Buffer input_buffer(std::begin(input_vector), std::end(input_vector), true);
cl::Buffer output_buffer(std::begin(output_vector), std::end(output_vector), false);

cl::EnqueueArgs enqueue_args(command_queue, cl::NDRange(5));

cl::KernelFunctor<cl::Buffer, cl::Buffer> functor(kernel);
functor(enqueue_args, input_buffer, output_buffer);

for (const auto& value : output_vector) {
    std::cout << value << ", ";
}

我希望在运行内核后打印输出向量缓冲区的结果，这应该等于input[n] + n，但是我只得到了填充输出向量的初始0s。我已经尝试了很多事情，但到目前为止还没有成功，所以为了清楚起见，我将其缩小了。内核确实构建了，运行它时没有任何错误，只是没有得到我希望的结果。我也看不到任何打印语句输出。

关于进一步的上下文，我的硬件最高支持 openCL 1.2，在 macOS 上运行，并且我已经定义了 openCL 定义以声明我正在使用 openCL 1.2。

谁能看到我在设置代码中做错了什么？

Answer 1

您的test_kernel 看起来不错。

但在 C++ 方面，您缺少一些东西：

1. 在设备上下文中创建缓冲区，以便 OpenCL 知道它应该在哪个设备的内存上分配内存：

   const int N = 5;
   cl::Buffer input_buffer(context, CL_MEM_READ_WRITE, N*sizeof(float));
   cl::Buffer output_buffer(context, CL_MEM_READ_WRITE, N*sizeof(float));
   

2. 您必须将Buffer 对象链接到Kernel 参数。否则，OpenCL 不知道内存中的哪些缓冲区对应test_kernel 的哪些参数。

   kernel.setArg(0, input_buffer);
   kernel.setArg(1, output_buffer);
   

3. 在执行内核之前，需要将input_vector从CPU内存复制到GPU内存中的input_buffer：

   command_queue.enqueueWriteBuffer(input_buffer, true, 0, N*sizeof(float), (void*)input_vector);
   

   注意：GPU 内存中的output_buffer 保持未初始化，并且可能包含随机值。由于您的test_kernel 将每个条目写入output_buffer 并且不读取output_buffer 的任何条目，因此此处不需要command_queue.enqueueWriteBuffer(output_buffer, ...)。
4. 执行kernel：

   const int local = 1; // GPU warp size is 32, so this should be 32 or a multiple of 32 to get full performance. For the test with N=5, I have set it to 1.
   const int global = ((N+local-1)/local)*local;
   cl::NDRange range_local = cl::NDRange(local);
   cl::NDRange range_global = cl::NDRange(global);
   command_queue.enqueueNDRangeKernel(kernel, cl::NullRange, range_global, range_local);
   command_queue.finish();
   

5. 将 output_buffer 从 GPU 内存复制回 CPU 内存中的 output_vector：

   command_queue.enqueueReadBuffer(output_buffer, true, 0, N*sizeof(float), (void*)output_vector);
   

   注意：enqueueReadBuffer 中的true 使其成为阻塞命令，表示在此行之后队列自动清空，数据传输完成；您不需要在此处添加额外的command_queue.finish();。
6. 更新的数据现在位于output_vector 的 CPU 内存中，可以在 CPU 上进一步处理或在控制台中打印。