如何使用 clEnqueueMigrateMemObjects 在相同上下文下的 GPU 之间传递缓冲区答案

【问题标题】：How to use clEnqueueMigrateMemObjects in order to pass buffers in between GPUs under a same context如何使用 clEnqueueMigrateMemObjects 在相同上下文下的 GPU 之间传递缓冲区
【发布时间】：2022-06-16 05:19:59
【问题描述】：

我正在尝试使用多个 GPU 来处理问题域。主要问题是我必须以某种方式找到一种在 GPU 之间有效传递缓冲区的方法。需要传递的缓冲区是每个 GPU 分配到的数组的边界值，因此一旦这些值在每个时间步更新，整个过程就可以在下一个时间步重复。

通过互联网搜索，我发现clEnqueueMigrateMemObects 就是为此目的。但是，我找不到任何有关跨 GPU 缓冲区传输的示例。我发现只有一种解释是this 帖子。我无法理解的部分是这部分（我放箭头的地方）

设备 1 上的命令队列：

迁移内存缓冲区1
将处理此缓冲区的内核排入队列
==> 保存与缓冲区 1 处理相关的最后一个事件 <==

设备 2 上的命令队列：

迁移内存缓冲区 1 - 使用队列 1 产生的事件来同步迁移。
将处理此缓冲区的内核排入队列

那么，示例代码如下所示？（假设我有两个 OpenCL 设备使用相同的平台和相同的上下文......）

...
cl_context context = clCreateContext(NULL, numDevices, devices, NULL, NULL, &status);
cl_command_queue cmdq_dev0, cmdq_dev1;

cmdq_dev0 = clCreateCommandQueue(context, devices[0], 0, &status);
cmdq_dev1 = clCreateCommandQueue(context, devices[1], 0, &status);

cl_mem dev0_buf, dev1_buf, common_buf;

dev0_buf = clCreateBuffer(context, CL_MEM_READ_WRITE, some_siz, NULL, &status);
dev1_buf = clCreateBuffer(context, CL_MEM_READ_WRITE, some_siz, NULL, &status);
common_buf = clCreateBuffer(context, CL_MEM_READ_WRITE, some_siz, NULL, &status);

status   =  clEnqueueWriteBuffer(cmdq_dev0, buf_arr  , CL_TRUE, 0, some_siz, dev0_arr, 0, NULL, NULL);
status   =  clEnqueueWriteBuffer(cmdq_dev0, common_buf, CL_TRUE, 0, common_siz, common_arr, 0, NULL, NULL);

status   =  clEnqueueWriteBuffer(cmdq_dev1, buf_arr  , CL_TRUE, 0, some_siz, dev1_arr, 0, NULL, NULL);
status   =  clEnqueueWriteBuffer(cmdq_dev1, common_buf, CL_TRUE, 0, common_siz, common_arr, 0, NULL, NULL);

/* build some opencl program */

cl_kernel kernel0, kernel1

kernel0 = clCreateKernel(program, \"kernel0\", &status);
kernel1 = clCreateKernel(program, \"kernel1\", &status);

status = clSetKernelArg(kernel0, 0, sizeof(cl_int), &dev0_arr  );
status = clSetKernelArg(kernel0, 1, sizeof(cl_int), &common_arr  );

status = clSetKernelArg(kernel1, 0, sizeof(cl_int), &dev1_arr  );
status = clSetKernelArg(kernel1, 1, sizeof(cl_int), &common_arr  );

/* part where kernels are executed */

status = clEnqueueNDRangeKernel(cmdq_dev0, kernel0, 3, NULL, something, NULL, 0, NULL, NULL);

status = clEnqueueMigrateMemObjects(cmdq_dev0, 1, &common_buf, CL_MIGRATE_MEM_OBJECT_HOST,0,NULL,NULL);

status = clEnqueueNDRangeKernel(cmdq_dev1, kernel0, 3, NULL, something, NULL, 0, NULL, NULL);

status = clEnqueueMigrateMemObjects(cmdq_dev1, 1, &common_buf, CL_MIGRATE_MEM_OBJECT_HOST,0,NULL,NULL);

...

此外，在将common_buf 缓冲区对象从设备0 传递到设备1 时，我应该在函数clEnqueueMigrateMemObjects 中指定命令队列，反之亦然。

谢谢。

标签： opencl

【解决方案1】：

您的代码中缺少的是使用事件对象的队列间同步（搜索 - OpenCL 事件，OpenCL 同步）。

使用多个队列时（无论是否在同一设备上），您必须在它们访问共享数据时同步它们的执行。排队的命令分离队列可以彼此并行运行。您加入不同队列的顺序无关紧要。它们彼此之间并没有按顺序执行，只对自己执行。

与您一样，典型场景是生产者-消费者类型的工作流。生产者队列生成稍后由消费者队列读取的数据。生产者队列必须为消费者队列提供一个事件对象，以便在读取生产数据之前等待。

我无法理解的部分是这部分（我放箭头的地方）

这部分意味着你应该从最后一次enqueue调用（clEnqueueNDRangeKernel的最后一个参数）中获取cl_event对象，并在等候名单另一个队列中clEnqueueMigrateMemObjects 的参数。

此外，在将 common_buf 缓冲区对象从设备 0 传递到设备 1 时，我应该在函数 clEnqueueMigrateMemObjects 中指定命令队列，反之亦然。

应该调用这个函数的队列是消费者打算将内核排入队列以访问缓冲区的队列。但是你应该只在另一个之后才这样做制片人queue 已完成对缓冲区的写入（通过使用前面描述的 cl_event 对象作为clEnqueueMigrateMemObjects 的等待列表）。

【讨论】：

谢谢你的小费！我从来没有意识到我需要传递一个cl_event 变量才能使clEnqueueMigrateMemObects() 正常工作。我有另一个问题。如果您有多个要迁移的内存对象，我应该如何在clEnqueueMigrateMemObects() 中指定？参考手册说“mem_objects”是内存对象的“列表”。还是应该为多个内存对象多次调用clEnqueueMigrateMemObects()？谢谢。
如果内存对象列表由同一个事件等待列表同步，则可以使用单个调用
嗯..我不太明白。因此，例如，如果我使用 4 个变量（比如说 a、b、c、d）执行 clEnqueueNDRangeKernel()，我可以使用 clEnqueueMigrateMemObjects() 通过一次调用传递整个 4 个变量吗？那么我的下一个问题是'mem_objects'应该是什么？会不会是clEnqueueMigrateMemObjects(cmdq_dev0,4,&a,&b,&c,&d,...) 之类的东西？
您应该提供一个 cl_mem 变量数组。请注意 clEnqueueMigrateMemObjects 接受一个指针（可能是一个数组或单个变量）
我明白那个。但我的困惑是，如果我想传递多个内存对象（数组数量）怎么办？