如何在Cuda中传递包含矩阵的结构体

Question

正如标题所说，我试图传递一个包含4个矩阵的结构体到Cuda内核。问题是我没有得到任何错误，但是每当我试图执行它时，程序就崩溃了。返回的所有值都是0，时钟值溢出。以下是我到目前为止所做的:

#define ROWS 700
#define COLS 1244
struct sobel {
int Gradient[ROWS][COLS];
int Image_input[ROWS][COLS];
int G_x[ROWS][COLS];
int G_y[ROWS][COLS];
};
__global__ void sobel(struct sobel* data)
{
int x = blockIdx.x * blockDim.x + threadIdx.x;
int y = blockIdx.y * blockDim.y + threadIdx.y;

int XLENGTH = ROWS;
int YLENGTH = COLS;
if ((x < XLENGTH) && (y < YLENGTH))
{
if (x == 0 || x == XLENGTH - 1 || y == 0 || y == YLENGTH - 1)
{
data->G_x[x][y] = data->G_y[x][y] = data->Gradient[x][y] = 0;
}
else
{
data->G_x[x][y] = data->Image_input[x + 1][y - 1]
+ 2 * data->Image_input[x + 1][y]
+ data->Image_input[x + 1][y + 1]
- data->Image_input[x - 1][y - 1]
- 2 * data->Image_input[x - 1][y]
- data->Image_input[x - 1][y + 1];
data->G_y[x][y] = data->Image_input[x - 1][y + 1]
+ 2 * data->Image_input[x][y + 1]
+ data->Image_input[x + 1][y + 1]
- data->Image_input[x - 1][y - 1]
- 2 * data->Image_input[x][y - 1]
- data->Image_input[x + 1][y - 1];
data->Gradient[x][y] = abs(data->G_x[x][y]) + abs(data->G_y[x][y]);
if (data->Gradient[x][y] > 255) {
data->Gradient[x][y] = 255;
}

}
}
}

int main() {
struct sobel* data = (struct sobel*)calloc(sizeof(*data), 1);
struct sobel* dev_data; 
cudaMalloc((void**)&dev_data, sizeof(*data));
cudaMemcpy(dev_data, data, sizeof(data), cudaMemcpyHostToDevice);
dim3 blocksize(16, 16);
dim3 gridsize;
gridsize.x = (ROWS + blocksize.x - 1) / blocksize.x;
gridsize.y = (COLS + blocksize.y - 1) / blocksize.y;
sobel <<< gridsize, blocksize >>> (dev_data);
cudaMemcpy(data, dev_data, sizeof(data), cudaMemcpyDeviceToHost);
free(data);
cudaFree(dev_data);
return 0;
}

我还必须为矩阵的每个obe分配设备内存吗?如有任何建议，不胜感激。

编辑:我在这里切换了几件事，但程序似乎忽略了嵌套的else语句，所有返回的值都是0。

Answer 1

你的代码中(至少)有2个错误。

1. 您没有为设备结构分配正确的大小:

   cudaMalloc((void**)&dev_data, sizeof(data));
   ^
   

   就像你在calloc呼叫中所做的那样，那应该是sizeof(*data)而不是sizeof(data)(cudaMemcpy呼叫可能也应该更新以反映此大小。)

2. 您需要在内核代码中进行适当的线程检查，如下所示:

   if (( x < XLENGTH ) && ( y < YLENGTH )){ // add this line
   if (x == 0 ||  x == XLENGTH - 1 || y == 0 || y == YLENGTH - 1)
   {
   data->G_x[x][y] = data->G_y[x][y] = data->Gradient[x][y] = 0;
   

   如果没有这个，下一个if测试行可能会允许越界线程参与归零操作。例如，x == 0将通过if-test的任何线程。但是那个线程可能有一个超出界限的y值。