为 OpenCL 内核分配了额外 32 字节的本地内存答案

【问题标题】：Extra 32 bytes of local memory allocated for OpenCL Kernel为 OpenCL 内核分配了额外 32 字节的本地内存
【发布时间】：2014-04-19 05:10:36
【问题描述】：

我想弄清楚为什么我正在尝试运行的 OpenCL 内核收到以下错误：

Context error: [CL_OUT_OF_RESOURCES] :
OpenCL Error : clEnqueueNDRangeKernel failed: local memory usage (16416 bytes) is more than available on the device (16384 bytes)

内核定义为：

__kernel void kernelFun(__read_only image2d_t src,
                        __global __write_only uchar8 *dst,
                        __global uchar4 *endpointBuffer,
                        __local uchar4 *pixelBuffer)
{
  ...
}

我正在使用标准的clSetKernelArg 例程分配本地内存：

clSetKernelArg(gKernel, 3, kPixelBufferBytes, NULL);

其中kPixelBufferBytes 等于 16384。

我的问题是，这些额外的 32 个字节是从哪里来的？

【问题讨论】：

顺便说一下，__write_only 和__read_only 属性仅适用于图像对象，对缓冲区没有任何意义。奇怪的是，实现不会为此抛出错误或警告。您在主机上创建缓冲区时为缓冲区指定只写/只读信息。

标签： opencl gpgpu gpu

【解决方案1】：

众所周知，某些 OpenCL 实现使用与本地内存相同的物理内存来存储内核参数。您有 32 个字节的内核参数，这可以解释这种差异的来源。

例如，NVIDIA GPU 肯定用于执行此操作（参见NVIDIA's original OpenCL best practices guide 的第 25 页）。

【讨论】：

这意味着每个指针是 8 个字节？这可能是因为我使用的是 64 位操作系统吗？
@Mokosha 是的，我假设是 64 位硬件。您可以通过使用 CL_DEVICE_ADDRESS_BITS 参数调用 clGetDeviceInfo 来检查您的 OpenCL 设备是 32 位还是 64 位。
CL_DEVICE_ADDRESS_BITS 的返回值为我的两个 OpenCL 设备 (GPU) 提供了 32。但是，我在 64 位机器上运行。这是一个错误吗？
有可能是GPU或者驱动只支持32位寻址，所以不一定是bug。您还可以编写一个运行sizeof(image2d_t) 和sizeof(global uchar4*) 的简单测试内核来找出这些参数有多大。也可能是内核参数在您的设备上是 8 字节对齐的。