理解和优化 pyCUDA 中的线程、块和网格

Question

我对 GPU 编程和 pyCUDA 还很陌生，我的知识还存在很大的差距。我花了很多时间搜索 SO，查看示例代码并阅读 CUDA/pyCUDA 的支持文档，但在解释中没有发现太多多样性，并且无法理解一些事情。

我无法正确定义块和网格维度。我目前运行的代码如下，目的是将数组 a 乘以浮点数 b:

from __future__ import division
import pycuda.gpuarray as gpuarray
import pycuda.driver as cuda
import pycuda.autoinit
from pycuda.compiler import SourceModule
import numpy as np

rows = 256
cols = 10
a = np.ones((rows, cols), dtype=np.float32)
a_gpu = cuda.mem_alloc(a.nbytes)
cuda.memcpy_htod(a_gpu, a)

b = np.float32(2)

mod = SourceModule("""
  __global__ void MatMult(float *a, float b)
  {
    const int i = threadIdx.x + blockDim.x * blockIdx.x;
    const int j = threadIdx.y + blockDim.y * blockIdx.y;
    int Idx = i + j*gridDim.x;
    a[Idx] *= b;
  }
  """)

func = mod.get_function("MatMult")

xBlock = np.int32(np.floor(1024/rows))
yBlock = np.int32(cols)

bdim = (xBlock, yBlock, 1)
dx, mx = divmod(rows, bdim[0])
dy, my = divmod(cols, bdim[1])
gdim = ( (dx + (mx>0)) * bdim[0], (dy + (my>0)) * bdim[1])
print "bdim=",bdim, ", gdim=", gdim

func(a_gpu, b, block=bdim, grid=gdim)
a_doubled = np.empty_like(a)
cuda.memcpy_dtoh(a_doubled, a_gpu)
print a_doubled - 2*a

代码应打印块尺寸 bdim 和网格尺寸 gdim，以及零数组。

这适用于较小的数组大小，例如，如果 rows=256 和 cols=10（如上例所示），输出如下：

bdim= (4, 10, 1) , gdim= (256, 10)
[[ 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.]]

但是，如果我增加 rows=512，我会得到以下输出：

bdim= (2, 10, 1) , gdim= (512, 10)
[[ 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.  2.  2. ...,  2.  2.  2.]
 [ 2.  2.  2. ...,  2.  2.  2.]]

表示对数组的某些元素进行了两次乘法运算。

但是，如果我将块尺寸强制设置为 bdim = (1,1,1)，问题将不再出现，对于较大的数组大小，我会得到以下（正确的）输出：

bdim= (1, 1, 1) , gdim= (512, 10)
[[ 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.]]

我不明白这个。这里发生了什么，这意味着随着数组大小的增加，这种定义块和网格维度的方法不再适用？此外，如果块的维度为 (1,1,1)，这是否意味着计算是按顺序执行的？

在此先感谢您的指点和帮助！

Answer 1

您在二维块的二维网格上操作。在您的内核中，您似乎假设 gridDim.x 会在网格的 x 维度中 return 线程数。

__global__ void MatMult(float *a, float b)
{
    const int i = threadIdx.x + blockDim.x * blockIdx.x;
    const int j = threadIdx.y + blockDim.y * blockIdx.y;
    int Idx = i + j*gridDim.x;
    a[Idx] *= b;
}

gridDim.xreturns块数rx网格方向，不是线程数。为了获得给定方向的线程数，您应该将一个块中的线程数乘以同一方向的网格中的块数：

int Idx = i + j * blockDim.x * gridDim.x

理解和优化 pyCUDA 中的线程、块和网格

Understanding and optimising threads, blocks, and grids in pyCUDA

python

parallel-processing

cuda

gpu

pycuda