Cuda/cudafy三维索引

Question

在没有掌握类似的堆栈溢出问题之后，我试着把我的头放在cuda上，然后我决定测试一个示例(我使用cudafy.net来实现c#，但是底层的cuda应该是可解析的)

我想做以下几件事。向内核发送一个4x4x4矩阵，并根据这个逻辑得到一个4x4x4：

if(input[x,y,z] == 1)
    output[x+1, y, z]++;

if(input[x,y,z] == 2)
    output[x-1, y, z]++;

我研究了下面这个愚蠢的例子。

public const int N = 1 * 1024;

//Omissions

gpu.Launch(128, 1, function, dev_a, dev_b, dev_c);

内核：

[Cudafy]
public static void add_0(GThread thread, int[] a, int[] b, int[] c)
{
    int tid = thread.blockIdx.x; // (tid 0 -> 127, from my understanding)
    while (tid < N)
    { 
        c[tid] = a[tid] + b[tid];
        tid += thread.gridDim.x;
    }
}

然后试图把它转到3d上。我不能把索引做好。说我有以下几点。(这里有三个数组来测试索引)

int size = 4;
int[] dev_delta = gpu.Allocate<int>(size * size * size);
int[] dev_space = gpu.Allocate<int>(size * size * size);
int[] dev_result = gpu.Allocate<int>(size * size * size);

gpu.Launch(new dim3(4, 4, 4), 1, "testIndex", dev_delta, dev_space, dev_result);

而内核：

[Cudafy]
public static void testIndex(GThread thread, int[] delta, int[] space, int[] result)
{
    int x = thread.blockIdx.x;
    int y = thread.blockIdx.y;
    int z = thread.blockIdx.z;
    delta[x]++;
    space[y]++;
    result[z]++;
}

我天真地预料到以下几点：

delta = {4,4,4,4,0,0,0,0,0, ... 0,0}
space = {4,4,4,4,0,0,0,0,0, ... 0,0}
result = {4,4,4,4,0,0,0,0,0 ... 0,0}

但我明白：

delta = {1,1,1,1,0,0,0,0,0, ... 0,0}
space = {1,1,1,1,0,0,0,0,0, ... 0,0}
result = {1,0,0,0,0,0,0,0,0 ... 0,0}

这对我来说毫无意义，显然我错过了什么。

问题：

我要启动多少个线程？

如何在三维(启动4x4x4线程并获取flat3DArrayx * sizeY * sizeZ +y* sizeZ +z的变量)中“索引”我的示例问题？

如何在二维中“索引”我的示例问题？(启动4x4线程，然后让每个线程处理长度为4的深度列)

我发现这可能是相关的Why is z always zero in CUDA kernel，如果这就是让我感到困惑的原因，我仍然很感激纯粹的-cuda回答来整理我的大脑。

Answer 1

我要启动多少个线程？您正在每个块启动一个线程，因此共有16个线程，因为没有使用Z参数。为了获得更好的性能，我建议也使用线程(至少128个线程，以及32个线程的倍数)。

如何在三维(启动4x4x4线程并获取flat3DArrayx * sizeY * sizeZ +y* sizeZ +z的变量)中“索引”我的示例问题？gpu.Launch方法的第二个参数是线程。因此，x、y和z可能分别为threadIdx.x、threadIdx.y和threadIdx.z。但是您也可能希望使用许多块，因此threadIdx.x + blockDim.x * blockIdx.x可能是一个很好的峰值。

您提供的here链接解释了Z维与此无关的原因。CUDAfy.Net公开了进一步调用库达运行时cuda / call调用的启动函数。在将参数从点网传递到本机环境时，CUDAfy.Net似乎忽略了Z参数，将其留给一个参数。(这很可能是因为数据自动化系统的早期版本不支持不同于一个的Z参数)。这个解释不是纯的，因为CUDA现在支持不同于一个的Z值，但是您的参数在CUDAfy.Net实现中被忽略了。