Roh*_*kan 1 python jit cuda numba
这是一个非常面向初学者的问题。我一直在研究常规 python 线程和 C 线程，并了解到我可以创建运行特定函数的线程，并且它们使用信号量和其他同步原语。
但是，我目前正在尝试使用 numba 的基于 python 的编译器来学习 Cuda。我编写了以下代码。
from numba import cuda
import numpy as np
@cuda.jit
def image_saturate(data):
pos_x, pos_y = cuda.grid(2)
if (pos_x, pos_y) <= data.shape:
data[pos_x, pos_y] = 1
if __name__ == "__main__":
image_quality = (128, 72)
image = np.zeros(image_quality)
thread_size = 32
block_size = image_quality
image_saturate[block_size, thread_size](image)
print(image)
但是，我觉得奇怪的是，我可以thread_size随心所欲地改变，结果是一样的——这意味着输出都是预期的。但是，当我改变的那一刻，block_size奇怪的事情开始发生，只有原始矩阵的大小被填充 - 所以它只是部分填充。
由此我了解到返回cuda.grid(2)块坐标。但是，我不应该能够获得实际的线程坐标和块坐标吗？
我是个新手，找不到任何在线学习资源。如果有人能回答我的问题并提供使用 Numba 学习 Cuda 的资源，那就太好了。
由此我了解到 cuda.grid(2) 返回块坐标。
事实并非如此。该语句返回完全限定的 2D 线程索引。返回值的范围将扩展到块坐标限制和线程坐标限制的乘积。
在 CUDA 中，内核启动的网格维度参数（您正在调用的block_size）以块数（每个方向）指定网格维度。内核启动的块维度参数（您正在调用的thread_size）指定网格中每个块的大小，以每个块的线程数为单位。
因此，启动的线程总数等于网格维度和块维度的乘积。总数将是所有这些事物在所有维度上的乘积。每个维度的总和将是该方向上的网格维度和该方向上的块维度的乘积。
因此，您的设计选择有问题，因为您有图像大小，并且将网格尺寸设置为等于图像大小。仅当每个块只有 1 个线程时，这才有意义。通过查看任何适当的 numba CUDA 代码（例如此处的代码） ，您会发现，一种典型的方法是将所需的总尺寸（在本例中为图像大小或维度）除以每个线程的线程数。块，以获得网格尺寸。
当我们这样做时，cuda.grid()内核代码中的语句将返回一个具有合理范围的元组。在您的情况下，它会将元组返回到正确从 x 中的 0..127 和 y 中的 0..71 开始的线程。您目前遇到的问题是该cuda.grid()语句可以返回 x 中 0..((128*32)-1) 范围内的元组，这是不必要的。
当然， if 语句的目标是防止越界索引，但是 的测试对<=我来说看起来并不正确。这是经典计算机科学中的 1 倍误差。shape应排除索引恰好与返回的限制相匹配的线程。
但是，当我更改 block_size 时，奇怪的事情开始发生，只有原始矩阵的大小才被填充 - 所以它只是部分填充。
确实不清楚您的期望是什么。您的内核设计是这样的：每个线程填充（最多）一个输出点。因此，合理的网格大小调整是将网格（x 中的总线程数和 y 中的总线程数）与图像尺寸相匹配。如果您遵循上述网格大小计算建议，然后将网格大小设置为小于图像大小，我预计输出图像的部分内容将不会被填充。不要那样做。或者，如果您必须这样做，请采用网格跨步循环内核设计。
说了这么多，以下是我将如何重写您的代码：
from numba import cuda
import numpy as np
@cuda.jit
def image_saturate(data):
pos_x, pos_y = cuda.grid(2)
if pos_x < data.shape[0] and pos_y < data.shape[1]:
data[pos_x, pos_y] = 1
if __name__ == "__main__":
image_x = 128
image_y = 72
image_quality = (image_x, image_y)
image = np.zeros(image_quality)
thread_x = 32
thread_y = 1
thread_size = (thread_x, thread_y)
block_size = ((image_x//thread_x) + 1, (image_y//thread_y) + 1) # "lazy" round-up
image_saturate[block_size, thread_size](image)
print(image)
它似乎对我来说运行正确。如果您现在建议您要做的是任意修改变量，block_size例如：
block_size = (5,5)
并且不进行其他更改，并期望输出图像完全填充，我想说这不是一个明智的期望。我不知道这怎么可能合理，所以我只想说 CUDA 不是这样工作的。如果您希望将数据大小与网格大小“解耦”，那么规范的方法就是使用已经讨论过的网格步幅循环。
我还删除了元组比较。我认为这与此无关。如果您仍然想使用元组比较，那么它应该与您基于 python 所期望的完全一样。CUDA 没有任何具体内容。