如何有效地交换OpenCL内存缓冲区？

Question

正如标题所示，我正在寻找如何有效地交换两个OpenCL缓冲区。我的内核使用两个全局缓冲区，一个作为输入，一个作为输出。但是，我在for循环中调用我的内核，每次设置内核参数、排队内核和交换缓冲区，因为前一个输出缓冲区将成为下一次迭代的输入缓冲区种子。

在这里，适当的方法是什么，以交换这两个缓冲区？我想象一下，将缓冲区复制回主机到已分配的数组之一，并使用clEnqueueWriteBuffer()和clEnqueueReadBuffer()将其复制到下一个输入缓冲区中，是一种低效的方法。否则，我只是使用一个临时的cl_mem变量来进行交换。

- voxeloctree

Answer 1

你不需要这样做，只需在第二次排队内核之前使用 clSetKernelArg 设置正确的内核参数即可。缓冲区将保留在设备上，不会复制任何东西回主机。

在这种情况下，你的缓冲区必须使用 CL_MEM_READ_WRITE 创建。

- Simon

---

我考虑过这个问题。但是似乎每次迭代都需要交替输入和输出内核参数，是吗？ - voxeloctree

是的 - 你会的。在每次迭代中交换并设置它们。这就是乒乓操作（http://www.mathematik.uni-dortmund.de/~goeddeke/gpgpu/tutorial.html#feedback2）。 - Ani

好的，我当然考虑过那个问题。但是如果我的问题没有表述清楚，那我很抱歉。“第二次”意味着两次，而不是数百万次。否则，大规模并行计算有什么意义呢？感谢你们两位。 - voxeloctree

您可以在调用 setArg 之前使用 std::swap(bufferA, bufferB);，以避免为偶数和奇数迭代使用单独的代码。 - Madcowswe

Answer 2

与前面的答案一样：不需要交换缓冲区。

然而，我不同意提出的答案。函数clSetKernelArg()不是线程安全的，也没有设计成在操作循环中调用。

正确的解决方案是创建两个使用相同程序和源代码创建的内核。这种方法更符合OpenCL编程理念“一个任务对应一个内核”。具有相同代码但不同参数的多个内核是可行的方式。

第一个内核将包括：

kernel1 = clCreateKernel(program, "mykernel", NULL);
clSetKernelArg(kernel1, 0, &buff1);
clSetKernelArg(kernel1, 1, &buff2);

另一个要翻译的内容是：

kernel2 = clCreateKernel(program, "mykernel", NULL);
clSetKernelArg(kernel2, 0, &buff2);
clSetKernelArg(kernel2, 1, &buff1);

这样，您就不需要在每次迭代中停止执行。您只需运行以下命令：

for(int it=0; it<iter; it++){
    clEnqueueNDRangeKernel(it%2 ? kernel1 : kernel2, ....);
}
clFinish(command);

这种方法肯定比更改内核参数更好，更高效，API调用更少。此外，在某些系统上，clSetKernelArgs()可能是一个阻塞调用，由于API实现不佳。因此，尽可能避免使用它们。

- DarkZeros

---

虽然我同意，clSetKernelArg不是线程安全的，但你关于“不允许你在完成上一个迭代之前运行下一个迭代”的说法是不正确的。一旦你调用了clEnqueueNDRangeKernel，所有参数都被复制了，所以你可以立即设置任何你想要的东西，在排队内核之后。 - alariq

我更新了我的回答。你是对的，99%的可用平台都没有阻止，只有一些旧的驱动程序或者实现不好的驱动程序在clSetKernelArgs()上被阻止。 - DarkZeros