内存对象 OpenCL 的基地址

Question

我想用 OpenCL 在 GPU 上遍历树，所以我在主机上将树组装在一个连续的块中，并更改所有指针的地址以便在设备上保持一致，如下所示：

TreeAddressDevice = (size_t)BaseAddressDevice + ((size_t)TreeAddressHost - (size_t)BaseAddressHost);

我想要内存缓冲区的基地址： 在主机上，我为缓冲区分配内存，如下所示： cl_mem tree_d = clCreateBuffer(...);

问题在于 cl_mems 是跟踪数据内部表示的对象。从技术上讲，它们是指向对象的指针，但它们不是指向数据的指针。从内核中访问 c​​l_mem 的唯一方法是通过 setKernelArgs 将其作为参数传入。

在这里http://www.proxya.net/browse.php?u=%3A%2F%2Fwww.khronos.org%2Fmessage_boards%2Fviewtopic.php%3Ff%3D37%26amp%3Bt%3D2900&b=28我找到了以下解决方案，但它不起作用：

__kernel void getPtr( __global void *ptr, __global void *out )

    {
    *out = ptr;
    }

可以如下调用

代码：

...

    cl_mem auxBuf = clCreateBuffer( context, CL_MEM_READ_WRITE, sizeof(void*), NULL, NULL );
    void *gpuPtr;

    clSetKernelArg( getterKernel, 0, sizeof(cl_mem), &myBuf );
    clSetKernelArg( getterKernel, 1, sizeof(cl_mem), &auxBuf );
    clEnqueueTask( commandQueue, getterKernel, 0, NULL, NULL );
    clEnqueueReadBuffer( commandQueue, auxBuf, CL_TRUE, 0, sizeof(void*), &gpuPtr, 0, NULL, NULL );

    clReleaseMemObject(auxBuf);

...

现在“gpuPtr”应该包含“myBuf”在GPU内存空间中的开头地址。

解决方案很明显，我找不到？创建缓冲区时如何取回指向设备内存的指针？

Answer 1

这是因为在 OpenCL 模型中，主机内存和设备内存是不相交的。设备内存中的指针在主机上没有任何意义。

您可以使用 clEnqueueMapBuffer 将设备缓冲区映射到主机内存。映射会将设备同步到主机，取消映射会将主机同步回设备。

更新。正如您在 cmets 中解释的那样，您希望将树结构发送到 GPU。一种解决方案是将所有树节点存储在一个数组中，用数组中的索引替换指向节点的指针。

Answer 2

正如 Eric 指出的，有两组内存需要考虑：主机内存和设备内存。基本上，OpenCL 试图通过在主机端的程序中引入缓冲区对象来隐藏这种交互的细节。现在，正如您所指出的，这种方法的问题在于，当我们想要做一些比 OpenCL 开发人员在其范围内预期或允许的更棘手的事情时，它会隐藏我们设备的细节。这里的解决方案是记住 OpenCL 内核使用 C99，并且该语言允许我们毫无问题地访问指针。考虑到这一点，我们可以要求将指针存储在一个无符号整数变量中以供以后引用。

您的实现在正确的轨道上，但它需要更多的 C 语法来完成传输。

OpenCL 内核：

// Kernel used to obtain pointer from target buffer
__kernel void mem_ptr(__global char * buffer, __global ulong * ptr)
{
    ptr[0] = &buffer[0];
}

// Kernel to demonstrate how to use that pointer again after we extract it.
__kernel void use_ptr(__global ulong * ptr)
{
    char * print_me = (char *)ptr[0];
    /* Code that uses all of our hard work */
    /* ... */
}

主持节目：

// Create the buffer that we want the device pointer from (target_buffer) 
//  and a place to store it (ptr_buffer).
cl_mem target_buffer = clCreateBuffer(context, CL_MEM_READ_WRITE, 
                                      MEM_SIZE * sizeof(char), NULL, &ret);
cl_mem ptr_buffer    = clCreateBuffer(context, CL_MEM_READ_WRITE,
                                      1 * sizeof(cl_ulong), NULL, &ret);

/* Setup the rest of our OpenCL program */    
/* .... */

// Setup our kernel arguments from the host...
ret = clSetKernelArg(kernel_mem_ptr, 0, sizeof(cl_mem), (void *)&target_buffer);
ret = clSetKernelArg(kernel_mem_ptr, 1, sizeof(cl_mem), (void *)&ptr_buffer);
ret = clEnqueueTask(command_queue, kernel_mem_ptr, 0, NULL, NULL);

// Now it's just a matter of storing the pointer where we want to use it for later.
ret = clEnqueueCopyBuffer(command_queue, ptr_buffer, dst_buffer, 0, 1 * sizeof(cl_ulong),
                          sizeof(cl_ulong), 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, ptr_buffer, CL_TRUE, 0,
                          1 * sizeof(cl_ulong), buffer_ptrs, 0, NULL, NULL);  

你有它。现在，请记住，您不必使用我使用的 char 变量；它适用于任何类型。但是，我建议使用 cl_ulong 来存储指针。对于可访问内存小于 4GB 的设备，这无关紧要。但对于地址空间较大的设备，则必须使用 cl_ulong。如果您绝对需要在设备上节省空间，但设备的内存 > 4GB，那么您可以创建一个结构，将地址的低 32 LSB 存储为 uint 类型，MSB 存储在小字体。