OpenCL ND-Range 边界？

Question

考虑一个执行向量加法的内核：

__kernel void vecAdd(__global double *a,
                     __global double *b,
                     __global double *c,
                     const unsigned int n)
{                                           
    //Get our global thread ID              
    int id = get_global_id(0);              

    //Make sure we do not go out of bounds  
    if (id < n)                             
        c[id] = a[id] + b[id];              
}

真的有必要将大小n 传递给函数，并检查边界吗？

我在n 上看到了相同的版本，但没有检查。哪一个是正确的？

更一般地说，我想知道如果要处理的数据大小与用户定义的 NR-Range 不同会发生什么。

是否会处理剩余的越界数据？

• 是这样，是怎么处理的？
• 如果不是，那是否意味着用户在编写内核时必须考虑边界？

OpenCL 是否指定了这些？

谢谢

Answer 1

这在主机代码指定工作组大小时很典型，因为在 OpenCL 1.x 中，全局大小必须是工作组大小的倍数。因此，如果您的数据大小为 1000，工作组大小为 128，则全局大小需要四舍五入为 1024。因此进行检查。在 OpenCL 2.0 中，此要求已被删除。

Answer 2

如果您不确定有多个 n 个工作项，则检查 n 是一个好主意。当您知道您只会使用 n 个工作项调用内核时，检查只会占用处理周期、内核大小和指令调度程序的注意力。

您传递给内核的额外数据不会发生任何事情。虽然如果您在某些时候不使用这些数据，您确实会浪费时间将其复制到设备上。

我喜欢让内核的工作组和全局大小独立于要完成的总工作量。在这种情况下，我需要传入“n”。

例如：

__kernel void vecAdd(  __global double *a, __global double *b, __global double *c, const unsigned int n)
{                                           
    //Get our global thread ID and global size
    int gid = get_global_id(0);              
    int gsize = get_global_size(0);              

    //check vs n using for-loop condition
    for(int i=gid; i<n; i+= gsize){
        c[i] = a[i] + b[i];              
    }
}

该示例将采用任意值 n 以及任何全局大小。每个工作项将处理每个第 n 个元素，从它自己的全局 id 开始。同样的想法也适用于工作组，由于内存局部性，有时性能优于我列出的全局版本。

如果您知道 n 的值是常数，通常最好对其进行硬编码（作为顶部的 DEFINE）。这将使编译器针对该特定值进行优化并消除额外的参数。此类内核的示例包括：DFT/FFT 处理、给定阶段的双调排序以及使用恒定维度的图像处理。