OpenCL使用寄存器值崩溃程序答案

【问题标题】：OpenCL usage of register value crashes programOpenCL使用寄存器值崩溃程序
【发布时间】：2016-07-06 06:05:34
【问题描述】：

我写完了一个用于热力学计算的 OpenCL 内核，并发现了一个非常奇怪的错误。我的内核是这样的：

__kernel void energy(... float3 dest, int nlocal, ...){
   int i = get_global_id(0);
   float3 ev = {0.0f, 0.0f, 0.0f};
   for(...){
      //some thermo calculations, adding values to evx and evy
       ev.x +=...;
       ev.y +=...;
    }

    //Then I want to save the result in dest[i].
    //Program exits at next two line
    dest[i].x = ev.x;
    dest[i].y = ev.y;

我收到“未映射的内存”和段错误错误。尝试使用 printf 打印值时出现相同的错误。似乎程序无法读取该值。写入它虽然有效！（可能是因为一些编译器优化）现在，如果我使用另一个浮点寄存器值，我会得到同样的错误。但是如果我把最后几行改成这样（不使用 ev.x 或 ev.y）

dest[i].x = i/nlocal*3.1f
dest[i].y = ...;

一切都按预期进行，我没有收到任何错误。

这也有效：

int i = ...
float3 = {0.0f, ...}
dest[i].x = ev.x;

但不知何故，在实际计算之后，它就不可能了。

该程序在 Nvidia K40m、Kepler 架构上运行。

【问题讨论】：

问题也可以在这里：ev.x +=...; ev.y +=...;。 CL 编译器将简单地删除不影响输出数组的代码。发布可重现的代码示例，否则我们无法帮助您。
感谢您的帮助，我已经更改了我的代码。它现在是整个内核。

标签： opencl gpu

【解决方案1】：

这在您的代码中看起来很可疑：

kernel(... __global int* neigh
   __global int* neighs = neigh+i;
   ...
   int j = neighs[k*n];
   ...

似乎您在neigh 中传递了一个指针数组，然后获取指针并使用它。 CL 中不允许使用指针，如果您传递指针，那么您将在 GPU 内存之外寻址，因此会崩溃。

也可能是您的向量没有正确计算，大小应该是：

res, nneigh = GLOBAL_SIZE
neighs = max(nneigh[])*n
x = max(neighs[])

还有可能您确实创建了比应有的更小的缓冲区（请记住，它们是浮点数和 float3，每个元素使用 32 位和 128 位）。 CL API 调用以字节定义（您应该使用sizeof()），而不是元素。

【讨论】：

我想我找到了导致段错误的问题。我通过 sizeof(numneigh) 两次，而不是 sizeof(neigh)。现在，当我使用 ev.x 或 ev.y 时出现 OutOfResources 错误......当我将其他值写入 res 时，一切都按预期进行。
似乎您正在传递一个指针数组，然后获取指针并使用它。 - 对我来说，这看起来更像是访问 2D 扁平数组。 @Noahnder 也值得包含主机代码，正如已经提到的，OpenCL 缓冲区的大小一定有问题。如果您将其他值写入 res 则其余代码将得到优化，因此您基本上不使用它并且问题消失了。
@doqtor 我认为使用neighs[k*n+i] 比获取指向元素i 的指针，然后在主要维度上再次偏移它更合乎逻辑。显得不自然。这就是为什么我猜测他可能在neighs 中有一个指针数组，而他这样做只是为了让它编译。我同意需要主机代码来调试此类问题。

【解决方案2】：

好的，我找到了答案，上面的代码正在运行。为了更好地理解，我更改了内核参数，并在此处发布代码时不自觉地纠正了错误。

int numneigh = nneigh[i] (stands for number of neighbors) is correct
in the original code I did this: 
int numneigh = neigh[i] (the neighbors)

感谢您的帮助，您猜测 neigh/nneigh 有问题是正确的，即使错误不在上面发布的代码中：P

【讨论】：