memory - 在 OpenCL(如 CUDA)中是否有供内核使用的主机内存的设备端指针？

Question

在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中共享内存的对应物)

Answer 1

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

它如何在 CUDA 上工作:

根据 this presentation和 official documentation .

关于 cudaHostGetDevicePointer 的报价:

Passes back device pointer of mapped host memory allocated by cudaHostAlloc or registered by cudaHostRegister.

CUDA cudaHostAlloc 与 cudaHostGetDevicePointer 的工作方式与 CL_MEM_ALLOC_HOST_PTR 与 MapBuffer 在 OpenCL 中的工作方式完全相同。基本上，如果它是一个独立的 GPU，结果会缓存在设备中，如果它是一个与主机共享内存的独立 GPU，它将直接使用内存。因此，在 CUDA 中不存在对离散 GPU 的实际“零复制”操作。

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

使用离散 GPU，指针指向 GPU 可以通过 DMA 直接传输内容的区域。否则，驱动程序会获取您的指针，将数据复制到 DMA 区域，然后启动传输。

但是在 OpenCL2.0 中，这显然是可能的，具体取决于您设备的功能。通过最精细的共享粒度，您可以使用随机分配的主机指针，甚至可以对主机使用原子操作，因此您甚至可以在主机运行时从主机动态控制内核。

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

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