PyCUDA 如何在启动内核时获取每个线程使用的寄存器数？

Question

我有一个内核，如何在启动内核时获取每个线程使用的寄存器数？我的意思是以 PyCuda 的方式。

一个简单的例子是：

__global__ 
void 
make_blobs(float* matrix, float2 *pts, int num_pts, float sigma, int rows, int cols) {
    int x = threadIdx.x + blockIdx.x * blockDim.x;
    int y = threadIdx.y + blockIdx.y * blockDim.y;
    if (x < cols && y < rows) {
        int  idx = y*cols + x;
        float temp = 0.f;
        for (int i = 0; i < num_pts; i++) {
            float x_0 = pts[i].x;
            float y_0 = pts[i].y;
            temp += exp(-(pow(x - x_0, 2) + pow(y - y_0, 2)) / (2 * sigma*sigma));
        }
        matrix[idx] = temp;
        }
    }

如果实际使用的数字超过最大值，有没有办法在不崩溃的情况下获取数字？

以上没问题，没有超过我机器的最大值。我只想以方便的方式获取号码。谢谢！

Answer 1

PyCuda 已经将其作为 Cuda 函数对象的一部分提供。属性称为 pycuda.driver.Function.num_regs。

下面是一个展示如何使用它的小例子：

import pycuda.autoinit
from pycuda.compiler import SourceModule

kernel_src = """
__global__ void 
make_blobs(float* matrix, float2 *pts, int num_pts, float sigma, int rows, int cols) {
int x = threadIdx.x + blockIdx.x * blockDim.x;
int y = threadIdx.y + blockIdx.y * blockDim.y;
if (x < cols && y < rows) {
    int  idx = y*cols + x;
    float temp = 0.f;
    for (int i = 0; i < num_pts; i++) {
        float x_0 = pts[i].x;
        float y_0 = pts[i].y;
        temp += exp(-(pow(x - x_0, 2) + pow(y - y_0, 2)) / (2 * sigma*sigma));
    }
    matrix[idx] = temp;
    }
}"""

compiledKernel = SourceModule(kernel_src)
make_blobs = compiledKernel.get_function("make_blobs")
print(make_blobs.num_regs)

请注意，您不需要使用 SourceModule。您还可以从例如加载模块一个 cubin 文件。可以在 documentation.

中找到更多详细信息

PyCUDA 如何在启动内核时获取每个线程使用的寄存器数？

PyCUDA how to get the number of used registers per thread when launching the kernels?

cuda

pycuda