1

在 OpenCL 中，我有 2 个工作组，每个工作组有 100 个工作项。所以我会做这样的事情：

....
clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&hDeviceMemInput);  
clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&hDeviceMemOutput); 
clSetKernelArg(kernel, 2, sizeof(cl_float) * 100, NULL);
clSetKernelArg(kernel, 3, sizeof(cl_int) * 1, &mCount);

clEnqueueNDRangeKernel(CmdQueue, Kernel, 1, 0, 200, 100, 0, 0, 0); 
....

OpenCL 代码：

__kernel square(
__global float *input,
__global float *output,
__local float *temp,
const unsigned int count)
{
int gtid = get_global_id(0);
int ltid = get_local_id(0);
if (gtid < count)
{
    temp[ltid] = input[gtid];
    output[gtid] =  temp[ltid] * temp[ltid];
}
}

据我了解，每个组都有一个 float[100] 本地临时变量。就我而言，设备上有两个 float[100]。如果有 n 个工作组，则设备上有 n 个 float[100]。那正确吗？__local float *temp 是否仅在设备上使用？我可以通过使用以下内容从内核中访问它：

clEnqueueReadBuffer(CmdQueue, ??, CL_TRUE, 0, 100* sizeof(cl_float),  
    host_temp, 0, 0, 0);

本地内存是否比全局内存快得多？你有使用本地内存的技巧吗？

4

1 回答 1

2

本地记忆是一种非常快的时间记忆。所以，不，你不能访问它或读回它。因为它被连续覆盖。事实上，内存并没有在设备中保留，因此您的 2 个工作组可能会（并且将会）使用相同的本地内存，但在不同的时间。如果您有 100 个组和 2 个计算单元......想象一下覆盖将发生多少次。

如果要读取本地内存的结果，必须先将其复制到全局，然后从那里读取。

本地内存的目的是在工作项之间共享一些东西，以获得时间中间结果和快速访问。之后它将被销毁。这对很多事情都很有用，一个简单的例子是过滤图像。

编辑：

您可以将本地内存视为一个寄存器，一个硬件资源。您不能将寄存器用作 RAM。就像您不能将本地内存用作全局内存一样。

于 2013-09-11T15:09:41.357 回答