在OpenCL中，是否有类似于CUDA中主机内存的设备端指针供内核使用？

Question

在OpenCL中，是否有类似于CUDA中主机内存的设备端指针供内核使用？

3

在CUDA中，我们可以通过主机内存的设备侧指针实现从主机内存到设备共享内存的内核管理数据传输。就像这样：

int  *a,*b,*c;          // host pointers
int *dev_a, *dev_b, *dev_c;     // device pointers to host memory

    …       

cudaHostGetDevicePointer(&dev_a, a, 0); // mem. copy to device not need now, but ptrs needed instead
cudaHostGetDevicePointer(&dev_b, b, 0);
cudaHostGetDevicePointer(&dev_c ,c, 0);

    …   

//kernel launch
add<<<B,T>>>(dev_a,dev_b,dev_c); 
// dev_a, dev_b, dev_c are passed into kernel for kernel accessing host memory directly.

在上面的例子中，内核代码可以通过 dev_a, dev_b 和 dev_c 访问主机内存。内核可以利用这些指针直接将数据从主机移动到共享内存中，而不需要通过全局内存进行中转。

但在OpenCL中似乎是不可能的？（在OpenCL中，本地内存是CUDA中共享内存的对应物）

- jxj

不可能拥有完全相同的功能。然而，这并不是核心特性。我想不出有哪种情况下你真的被迫执行这种操作。难道在调用内核之前，不能只调用clEnqueueWriteBuffer()吗？设备本地缓存对你来说是一个限制因素吗？ - DarkZeros

这是一种优秀的方式，可以在传输和计算重叠时使用。避免了从主机端发出显式复制操作，该操作只能将数据从主机复制到设备全局内存。通过设备端指针，可以直接在主机和共享内存之间传输数据。它允许设备根据需要安排计算和数据传输，这意味着数据传输可能会被隐藏。传统的方法是多流（CUDA）和多命令队列（OpenCL）。传统的方法需要在主机端进行显式调度，这使得整体代码有些混乱/臃肿。 - jxj

1个回答

网页内容由stack overflow 提供, 点击上面的

可以查看英文原文，
原文链接

- sharpneli · Accepted Answer

您可以在OpenCL中找到完全相同的API。

CUDA的工作原理：

根据this presentation和official documentation。

关于cudaHostGetDevicePointer的重要引用：

传递由cudaHostAlloc分配或由cudaHostRegister注册的映射主机内存的设备指针。

CUDA cudaHostAlloc与MapBuffer在OpenCL中的CL_MEM_ALLOC_HOST_PTR完全相同。基本上，如果它是离散GPU，则结果将被缓存在设备中，如果它是具有与主机共享内存的离散GPU，则会直接使用内存。因此，在CUDA中不存在实际的“零拷贝”操作。

函数cudaHostGetDevicePointer不接受原始的malloc指针，就像在OpenCL中的限制一样。从API用户的角度来看，这两种方法完全相同，允许实现进行几乎相同的优化。

对于独立GPU，您获得的指针指向一个区域，该区域能够通过DMA直接传输数据。否则，驱动程序将获取您的指针，将数据复制到DMA区域，然后启动传输。

但是，在OpenCL2.0中，这是明确可能的，具体取决于您的设备能力。通过最细粒度共享，您可以使用随机malloc的主机指针，并且甚至可以与主机使用原子操作，因此您甚至可以在内核正在运行时动态控制主机。

http://www.khronos.org/registry/cl/specs/opencl-2.0.pdf

请参阅第162页的共享虚拟内存规范。请注意，即使在编写内核时，这些仍然只是从内核视角看到的__global指针。