与opencl的累积和

Question

我尝试用opencl实现累积和如下：

__kernel void cumsum(__global float *a)
{
    int gid = get_global_id(0);
    int n = get_global_size(0);

    for (int i = 1; i < n; i <<= 1)
        if (gid & i)
            a[gid] += a[(gid & -i) - 1];
}

我使用 pyopencl 调用了这段代码：

import pyopencl as cl
import pyopencl.array as cl_array
import numpy as np

a = np.random.rand(50000).astype(np.float32)

ctx = cl.create_some_context()
queue = cl.CommandQueue(ctx)

a_dev = cl_array.to_device(queue, a)

with open("imm/cluster.cl", 'r') as f:
    prg = cl.Program(ctx, f.read()).build()

prg.cumsum(queue, a.shape, None, a_dev.data)
print(np.cumsum(a)[:33], a_dev[:33])

但是，前 32 个数字是正确的，之后它们是错误的（太低）。这与工作组规模有关吗？我该如何解决这个问题？

Answer 1

当 i 变得足够大时，您将读取另一个工作组的输出。 OpenCL 执行模型中的任何内容都不能保证其他工作组将完成执行。

通常情况并非如此，您将读取部分总和，最后得到的值低于预期。