OpenCL - 4D 数组上的元素操作

如何解决OpenCL - 4D 数组上的元素操作

我正在尝试编写 OpenCL 代码来对多维数组执行元素操作。

我知道 OpenCL 缓冲区是扁平化的，这使得索引有点棘手。我在处理二维数组时成功了，但对于 3+ 维数组，我要么出现索引错误，要么出现错误结果。

更令人惊讶的是，我使用了与 2D 情况相同的索引原理/公式。

二维案例：

__kernel void test1(__global int* a,__global int* b,__global int* c,const int height) {
    int i = get_global_id(0);
    int j = get_global_id(1);
    c[i + height * j] = a[i + height * j] + b[i + height * j];
}

正确。

3D 案例：

__kernel void test1(__global int* a,const int dim1,const int dim2) {
    int i = get_global_id(0);
    int j = get_global_id(1);
    int k = get_global_id(2);

    int idx = i + dim1 * j + dim1 * dim2 * k;
    c[idx] = a[idx] + b[idx];
}

错误的结果（通常是输出缓冲区填充了非常接近 0 的值）。

4D 案例：

__kernel void test1(__global int* a,const int dim2,const int dim3) {
    int i = get_global_id(0);
    int j = get_global_id(1);
    int k = get_global_id(2);
    int l = get_global_id(3);

    int idx = i + dim1 * j + dim1 * dim2 * k + l * dim1 * dim2 * dim3;
    c[idx] = a[idx] + b[idx];
}

这是索引错误：enqueue_knl_test1 pyopencl._cl.LogicError: clEnqueueNDRangeKernel failed: INVALID_WORK_DIMENSION

解决方法

在 4D 情况下，您只是错误地使用了 API。 OpenCL 不支持无限数量的全局/局部维度。最多 3 个。

在二维情况下，您的索引似乎是错误的。假设 row-major 数组。应该是 i + j * width 而不是 i + j * height。

在 3D 情况下，内核内部的索引似乎没问题，假设内存布局为行优先，并且 dim1 等于 cols（宽度），dim2 等于行（高度）。但无论如何，你的问题缺乏背景：

• 输入缓冲区分配和初始化。
• 内核调用代码（参数、工作组和全局大小）。
• 结果收集。同步。
• 您可能正在访问超出分配的缓冲区大小。应该检查一下。

错误地执行这些步骤很容易导致意想不到的结果。即使你的内核代码没问题。

如果您想调试索引问题，最简单的方法是编写一个简单的内核来输出计算出的索引。

__kernel void test1(__global int* c,const int dim1,const int dim2) {
    int i = get_global_id(0);
    int j = get_global_id(1);
    int k = get_global_id(2);

    int idx = i + dim1 * j + dim1 * dim2 * k;
    c[idx] = idx;
}

然后您应该期望值线性增加的结果。我会从一个工作组开始，然后继续使用多个工作组。

此外，如果您在数组之间执行简单的逐元素操作，那么使用一维索引要简单得多。您可以简单地使用一维工作组和等于元素数量的全局大小（四舍五入以适合工作组暗淡）：

__kernel void test1(__global int* a,__global int* b,__global int* c,const int total) {
    // no need for complex indexing for elementwise operations
    int idx = get_global_id(0);
    if (idx < total)
    {
       c[idx] = a[idx] + b[idx];
    }
}

您可能会将 local_work_size 设置为硬件允许的最大大小（例如，Nvidia 为 512，AMD 为 256），并将 global_work_size 设置为四舍五入为 {{1} 倍数的元素总数}.见clEnqueueNDRangeKernel。

2D & 3D dims 通常用于访问 2D / 3D 空间中的相邻元素的操作。比如图像卷积。