更改 CUDA 中多个单元格的值

Question

它必须是一个简单的，虽然我找不到答案。我正在编写一个必须计算元胞自动机状态的程序，为了了解 CUDA 的工作原理，我首先尝试编写一个非常简单的程序。它需要一个矩阵，每个线程都必须在其单元格以及该单元格上方和下方的单元格中增加一个值。所以，如果我给它下面的矩阵：

[0 0 0 0 0 0 0]
[0 0 0 0 0 0 0]
[0 0 0 0 0 0 0]
[0 0 0 0 0 0 0]
[0 0 0 0 0 0 0]
[0 0 0 0 0 0 0]
[0 0 0 0 0 0 0]

我希望得到以下结果：

[2 2 2 2 2 2 2]
[3 3 3 3 3 3 3]
[3 3 3 3 3 3 3]
[3 3 3 3 3 3 3]
[3 3 3 3 3 3 3]
[3 3 3 3 3 3 3]
[2 2 2 2 2 2 2]  

第一行的值为 2，因为它上面没有可以将第一行的值再增加一次的行。并且以类似的方式，最后一行的值为 2。
但我得到一个如下所示的矩阵：

[2 2 2 2 2 2 2]
[3 3 3 3 3 3 3]
[3 3 3 3 3 3 3]
[3 3 3 3 2 2 2]
[2 2 2 2 2 2 2]
[2 2 2 2 3 3 3]
[2 2 2 2 2 2 2]  

我不明白为什么第 4、5 和第 6 行的值为 2 - 必须是 3，而不是 2。
这是我的代码：

import numpy
import pycuda.autoinit
import pycuda.driver as cuda

from pycuda.compiler import SourceModule

w = 7

mod = SourceModule("""
        __global__ void diffusion(  int* result, int width, int height) {

            int xIndex = blockDim.x * blockIdx.x + threadIdx.x;
            int yIndex = blockDim.y * blockIdx.y + threadIdx.y;

            int flatIndex = xIndex + width * yIndex;
            int topIndex = xIndex + width * (yIndex - 1);
            int bottomIndex = xIndex + width * (yIndex + 1);

            int inc = 1;

            result[flatIndex] += inc;

            result[bottomIndex] += inc;

            result[topIndex] += inc;
        }

        """)

diff_func   = mod.get_function("diffusion")


def diffusion(res):

    height, width = numpy.int32(len(res)), numpy.int32(len(res[0]))

    diff_func(
        cuda.InOut(res),
        width,
        height,
        block=(w,w,1)
        )

def run(res, step):

    diffusion(res)
    print res

res   = numpy.array([[0 \
                        for _ in xrange(0, w)]\
                        for _ in xrange(0, w)], dtype='int32')

run(res, 0)  

还有一件更有趣的事：如果我评论以下行之一：

result[bottomIndex] += inc;
result[topIndex] += inc;  

一切都按预期工作，没有任何意外值。在某些情况下，CUDA 似乎无法在一个线程中处理三个相邻单元格的值。

Answer 1

您有所谓的内存竞争：多个独立线程试图同时更新内存中的相同值。 CUDA 内存模型没有定义当两个线程尝试同时更新同一个内存位置时会发生什么。

解决方案是使用原子内存操作（有关更多信息，请参阅 CUDA 编程指南），或者使用不同的方法更新相邻单元格（例如，为网格着色并像在网格中的不同通道上的彩色单元格一样更新） .