（Pyopencl 同时修改所有线程的值

Question

很简单，我有以下修改C[0]值的内核，其中C是一个只有一个元素的数组。

__kernel void sigma(__global float *A, __global float *B, __global float *C) {
    int i = get_global_id(0);
    printf("Adding %.2f + %.2f", A[i], B[i]);
    C[0] += A[i] + B[i];
}

问题是，最后C[0]有最后完成的线程的值，具体在这个例子中我得到以下

Adding 1.00 + 0.00
Adding 2.00 + 1.00
Adding 3.00 + 1.00
Adding 4.00 + 1.00
[5.]

最后C[0]是4.00 + 1.00。我想要的是 C[0] 成为 (1.00 + 0.00) + (2.00 + 1.00) + (3.00 + 1.00) + (4.00 + 1.00)。所以我想将每个线程的 A[i] 和 B[i] 添加到 C[0].

此外，我不只是在寻找补充，我希望它与任何功能或操作兼容。

这可能是多余的但是在主机代码中我只是做了最低限度的工作以将数据传递给内核。这个问题与主机代码有关吗？

import pyopencl as cl, numpy as np; 
ctx = cl.create_some_context(); queue = cl.CommandQueue(ctx); mf = cl.mem_flags
M = np.array([1, 2, 3, 4]).astype(np.float32) # A
V = np.array([0, 1, 1, 1]).astype(np.float32) # B
a = np.array([0]).astype(np.float32) # C
# Transfer data to GPU
A_GPU = cl.Buffer(ctx, mf.READ_ONLY | mf.COPY_HOST_PTR, hostbuf=M) 
B_GPU = cl.Buffer(ctx, mf.READ_ONLY | mf.COPY_HOST_PTR, hostbuf=V)
C_GPU = cl.Buffer(ctx, mf.WRITE_ONLY, a.nbytes)
c = np.zeros(shape=a.shape, dtype= np.float32) # array to copy the result
kernel.sigma(queue, M.shape, None, A_GPU, B_GPU, C_GPU)
cl.enqueue_copy(queue, c, C_GPU).wait()

Answer 1

您的内核编写方式效率低下，在许多情况下会导致不确定的结果。通常，您应该避免让多个线程（工作项）同时写入同一地址。在你的情况下 C[0] += A[i] + B[i]; 是一个 read modify write 序列，从并行访问的角度来看，这使事情变得更糟。

您需要使用的是平行缩减模式。此模式允许您有效且确定地将数组中的所有元素相加为单个值（或者在您的情况下 - 2 个数组）。它还支持除加法之外的许多其他关联运算符。

网上有很多资源，您可以根据自己的情况进行调整。

Answer 2

正如@Elad Maimoni 所说，在 PyOpenCl 中执行缩减的一种方法是使用版本 2.0 中可用的函数 work_group_reduce_add。

在python中可以这样实现

内核

kernel = cl.Program(ctx, """
__kernel void resum(__global float *A, __global float *B, __global float *a) {
    int i = get_global_id(0);
    a[0] = work_group_reduce_add(A[i] + B[i]);
}
""").build(options='-cl-std=CL2.0') # Build using cl 2.0

主机代码的其余部分

# Some example arrays
a = np.array([1, 2, 3, 4, 5, 6, 7, 8, 9])
b = np.array([9, 8, 7, 6, 5, 4, 3, 2, 1])
c = np.array([0])
d = np.array([0])

# Create GPU data
a = cl.Buffer(ctx, mf.READ_ONLY | mf.COPY_HOST_PTR, hostbuf=a)
b = cl.Buffer(ctx, mf.READ_ONLY | mf.COPY_HOST_PTR, hostbuf=b)
c = cl.Buffer(ctx, mf.READ_ONLY | mf.COPY_HOST_PTR, hostbuf=c)

kernel.resum(queue, (10, ), None, a, b, c)
cl.enqueue_copy(queue, d, c)
print(d)

这将输出 [90].

（Pyopencl 同时修改所有线程的值

(Py)OpenCL modify value from all threads simultanously

python

opencl

pyopencl