这是在设备端分配内存的语句 safeCall(cudaMallocPitch((void *)&d_data, (size_t)&pitch, (size_t)(sizeof(float)*width), (size_t)height));
这是拷贝数据的语句，safeCall(cudaMemcpy(d_data, h_data, sizeof(float)pitchheight, cudaMemcpyHostToDevice));
运行时报错，说内存访问冲突，是不是发生了越界呢？
楼主您好，您不能简单的使用cudaMemcpy的：
您原始的host data大小为width * height(这里指字节，下同）， 但您分配的目标设备缓冲区大小为pitch * height的，
其中pitch将是大于等于width的一个不固定的值（和设备有关）。
因为pitch >= width, 所以pitch * height >= width * height的。
所以如果直接使用cudaMemcpy, 只要您的width不是很凑巧，均会导致源长度不够的。
解决方案：
请改用cudaMemcpy2D(目标缓冲区, 目标pitch, 源缓冲区, 源pitch, …)来完成您的复制（注意单位均为字节）
感谢深夜来访。
非常感谢您的回答，但是我改成safeCall(cudaMemcpy2D(d_data,(size_t)pitch, h_data,(size_t)width,(size_t)(sizeof(float)*width),(size_t)height, cudaMemcpyHostToDevice));之后，运行时还是报错“invalid pitch argument”，纠结
而且我还发现，在safeCall(cudaMallocPitch((void *)&d_data, (size_t)&pitch, (size_t)(sizeof(float)*width), (size_t)height));之后，pitch变为原来的4倍，在我输入的这个矩阵中，pitch的初始值刚好等于width，我知道cudamallocpitch返回的pitch值一般大于或等于width，但是我通过输出发现，返回的pitch刚好等于原来的4倍，这是为什么呢？我在这个语句之后加了pitch /= sizeof(float);但是执行cudaMemcpy2D时又发生上述的错误，调了很久，搞不懂，求指教
显然楼主您没有认真看我之前的回复：
cudaMemcpy2D(d_data,(size_t)pitch, h_data,(size_t)width,(size_t)(sizeof(float)*width),(size_t)height, cudaMemcpyHostToDevice)
请将第3个参数从width改成width * sizeof(float)。
前文已经要求过您使用字节单位了。
关于为何您的pitch(字节单位)是width(元素单位)的4倍, 那是因为您的width转换成字节单位后正好是设备要求的pitch或者它的整数倍。
（这意味着您原来的width转换为字节后是512B的倍数一般，或者说，width（元素单位）是128的倍数）
（默认pitch是512N的形式）
版主，您好！我在cudamallocpitch后加上了pitch /= sizeof(float);所以pitch又变为了元素单位，之后我将这条语句的参数都改成字节单位，改成了safeCall(cudaMemcpy2D(d_data,(size_t)(sizeof(float)*pitch), h_data,(size_t)(sizeof(float)*width),(size_t)(sizeof(float)*width),(size_t)height, cudaMemcpyHostToDevice)); 结果又出现访问内存的错误
那只能有2个可能，
（1）是您的源缓冲区地址无效，或者地址有效，但里面至少连width * height * 4字节都没有。
（2）则是您的错误发生在其他位置，例如发生在您调用cudaMemcpy2D()之前（例如您试图定义>1MB的局部数组，此时debug下会立刻触发错误，根本就没执行对cudaMemcpy2D的调用）。
因为您没有给出过完整代码，还请您检查具体是哪一种。