CUDA - memcpy2d - 错误间距

Question

我刚启动CUDA编程，并试图执行下面显示的代码。这个想法是复制一个二维数组到设备中，计算所有元素的总和，然后检索总和（我知道这个算法没有并行化，实际上它做了更多的工作，然后是必要的。作为memcopy的练习）。

#include<stdio.h> 
#include<cuda.h> 
#include <iostream> 
#include <cutil_inline.h> 

#define height 50 
#define width 50 

using namespace std; 

// Device code 
__global__ void kernel(float* devPtr, int pitch,int* sum) 
{ 
int tempsum = 0;  
for (int r = 0; r < height; ++r) { 
     int* row = (int*)((char*)devPtr + r * pitch); 
     for (int c = 0; c < width; ++c) { 
      int element = row[c]; 
      tempsum = tempsum + element; 
     } 
    } 
*sum = tempsum; 
} 

//Host Code 
int main() 
{ 

int testarray[2][8] = {{4,4,4,4,4,4,4,4},{4,4,4,4,4,4,4,4}}; 
int* sum =0; 
int* sumhost = 0; 
sumhost = (int*)malloc(sizeof(int)); 

cout << *sumhost << endl; 

float* devPtr; 
size_t pitch; 
cudaMallocPitch((void**)&devPtr, &pitch, width * sizeof(int), height); 
cudaMemcpy2D(devPtr,pitch,testarray,0,8* sizeof(int),4,cudaMemcpyHostToDevice); 

cudaMalloc((void**)&sum, sizeof(int)); 
kernel<<<1, 4>>>(devPtr, pitch, sum); 
cutilCheckMsg("kernel launch failure"); 
cudaMemcpy(sumhost, sum, sizeof(int), cudaMemcpyDeviceToHost); 

cout << *sumhost << endl; 

return 0; 
} 

此代码编译得很好（在4.0 sdk发布候选版本上）。但是当我尝试执行，我得到

0 
cpexample.cu(43) : cutilCheckMsg() CUTIL CUDA error : kernel launch failure : invalid pitch argument. 

这是不幸的，因为我不知道如何解决它;-(。据我所知，在球场是在内存偏移允许更快的复制的数据。然而，这样的间距在设备存储器中仅使用，而不是在主机存储器，是吗？因此，我的主机存储器的间距应为0，不应该吗？

此外我还想问另外两个问题：

• 如果我声明一个像int * sumhost这样的变量（见上面），这个地方在哪里呃指向？首先到主机内存和cudaMalloc到设备内存之后？
• cutilCheckMsg在这种情况下非常方便。是否有类似的调试功能，我应该知道？

Answer 1

在这一行代码的：

cudaMemcpy2D(devPtr,pitch,testarray,0,8* sizeof(int),4,cudaMemcpyHostToDevice); 

你说的 testarray源间距值等于 0

，但是这怎么可能当间距公式为T* elem = (T*)((char*)base_address + row * pitch) + column？如果我们用0替代该公式中的音高值，那么当在某个2维（x，y）有序对偏移量处查找地址时，我们将无法获得正确的值。需要考虑的一件事是音高值的规则是pitch = width + padding。在主机上，填充通常等于0，但宽度不是0，除非阵列中没有任何内容。在硬件方面可能会有额外的填充，这就是为什么pitch的值可能不等于数组声明的宽度。因此，您可以根据填充值得出pitch >= width的结论。所以即使在主机端，源间距的值也应该至少是每个字节的大小，这意味着在testarray的情况下，它应该是8*sizeof(int)。最后，主机中2D阵列的高度也只有2行，而不是4。

作为对分配指针会发生什么问题的回答，如果您为malloc()分配一个指针，则该指针被赋予一个驻留在主机内存中的地址值。所以你可以在主机端解除引用，但不能在设备端引用。另一方面，分配给cudaMalloc()的指针被赋予指向驻留在设备上的存储器的指针。因此，如果您在主机上取消引用它，它不会指向主机上分配的内存，并且会出现不可预知的结果。尽管将该指针地址传递给设备上的内核也是可以的，因为当它在设备端取消引用时，它指向设备本地可访问的内存。总的来说，CUDA运行时间将这两个存储位置分开，提供将在设备和主机之间来回复制的存储器复制功能，并根据所需的方向使用来自这些指针的地址值作为副本的来源和/或目的地（主机到设备或设备到主机）。现在，如果你采取了同样的int*，先用malloc()分配，然后（后上的指针调用希望free()）与cudaMalloc()，你的指针将首先有一个指向主机内存中，然后设备存储器的地址。你将不得不跟踪它在，为了避免非关联化，这是设备或主机上取决于它在主机代码或设备代码是否被解除引用地址不可预知的结果状态。