OpenCL 中的并发内核

Question

我想知道如何同时并行执行两个或多个不同的内核？显然在使用 OpenCL 的同一个 GPU 中。我的主要想法是使用两个不同的内核（内核 A 和内核 B），但它们需要使用相同的内存（我不想通过为“a”和“b”指针中的每个指针使用一个缓冲区来复制内存） .那么有没有另一种方法可以使用高效的内存技术来完成双重执行？内核的代码如下：内核 A：

_kernel  void kernelA(global struct VectorStruct* a, int aLen0, global struct VectorStruct* b, int bLen0, global struct VectorStruct* c, int cLen0) {
int i = get_local_id(0);
c[(i)].x = a[(i)].x + b[(i)].x; }

内核 B：

_kernel  void kernelB(global struct VectorStruct* a, int aLen0, global struct VectorStruct* b, int bLen0, global struct VectorStruct* d, int cLen0){ int i = get_local_id(0); d[(i)].y = a[(i)].y + b[(i)].y; }

结构体 VectorStruct 的定义如下：

struct VectorStruct { int x; int y; };

在主机代码中我必须创建四个指针：向量结构*一个向量结构* b 矢量结构* c 向量结构* d 指针“a”和“b”具有我将传输到 GPU 的数据。指针c存放内核A的结果，指针d存放内核B的结果

Answer 1

您可以在并发命令队列中使用 clEnqueueNDRangeKernel() 将您的 2 个内核排入队列，即 CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE 在 clCreateCommandQueue 期间传递的队列.然后将两个创建的事件对象传递给缓冲区读取或映射调用，以从主机读取结果。请注意，并非所有硬件和 OpenCL 实现都支持不同内核的并发执行，因此它们最终可能会在某种程度上被序列化。

您还可以使用多个串行命令队列实现类似的效果。

对于您的简单内核，最好使用 float2 来表示您的向量并在单个内核中执行向量化 (SIMD) 加法。 OpenCL 编译器应选择向量运算并自动将运算分布到并行硬件上。

对于稍微复杂一些的操作效果不佳的情况，您可以将向量的 x 和 y 坐标表示为一个 2 元素数组，并简单地在一个有效的内核上将两倍数量的工作项加入队列在交替维度上。

这两种方法都会为您提供更高效的内存访问模式。

请注意，您对 get_local_id(0) 的使用可能有误，具体取决于您想要实现的目标 - 在这种情况下您可能希望使用 get_global_id(0)。

OpenCL 中的并发内核

concurrent kernel in OpenCL

memory

kernel

opencl