CUDA 资料汇总

CUDA 资料汇总 环境 本机环境

10.2和11.1两个版本，为了适配对应的pytorch，因此在使用前应当注意将环境变量更换到对应版本。

常用指令 常用代码 编程技巧 常用代码规范

1. 尽管是void但是还是可以使用return，来保证不出现报错if(Idx >= n) return; //超过数组大小，直接返回

基础知识 cuda修饰符

1. __host__函数可以直接调用__host__函数，但不能直接调用__device__函数；__host__函数可以通过传递运行时参数来调用__global__函数，同样也不能像调用__host__函数那样直接调用。而能调用__device__函数的只有__global__函数或者__device__函数
2. __device__和__global__函数也可以继续通过传递运行时参数调用__global__函数，实现二级甚至二级以上的并行。用拓扑结构图来表示则是：3. device，__shared__与__constant__修饰变量

资源

1. cuda安装目录下有DOC文件夹，可以找到官方文档！但是不知道为啥加载总是很慢，我已经把他们放在了固态啊。
2. 一个中文手册，一个中文书。
3. cuda_sample
4. https://blog.csdn.net/shanglianlm/article/details/90108295
5. 使用nsight和trace可以可视化运算路径
6. https://zhuanlan.zhihu.com/p/34587739
7. 基础函数、cudaerror_t报错https://zhuanlan.zhihu.com/p/146691161

报错

实际编程时请将每一个cuda函数都加上判断返回值的异常处理体系，否则调试量将会十分显著地上升！

(ide编译失败)Compile Error 环境原因

1. 使用随机数生成器时候，需要注意的是，用VS直接创建的cuda工程，直接编译上述代码很可能出现一些编译错误：
   error LNK2019: 无法解析的外部符号 curandCreateGenerator，函数 main 中引用了该符号
   error LNK2019: 无法解析的外部符号 curandGenerate，函数 main 中引用了该符号
   有编程经验的朋友们看到LNK就知道这是链接器报错，肯定是少链接了一些文件。没错，只需要在项目-属性-链接器-输入-附加依赖项中添加一项"curand.lib"即可。

2. 还有可能是报错的cu文件并没有用CUDA C/C++方法被编译（但与后缀无关，“.cu”文件后缀名完全可以改为“.cpp”，只要保持使用nvcc编译器）：

3. error LNK2005: 函数名 已经在 xxx.cu.obj 中定义。这个错误往往会伴随“fatal error LNK1169: 找到一个或多个多重定义的符号”出现。
   

4. CPP文件不能包含.cu文件

5. 打开sample时，错误 : 项目“C:ProgramDataNVIDIA CorporationCUDA Samplesv10.2_SimplesimpleCudaGraphssimpleCudaGraphs_vs2019.vcxproj”配置“Debug|x64”的 Designtime 生成失败。IntelliSense 可能不可用。
   设置环境变量 TRACEDESIGNTIME = true 并重启 Visual Studio 以进行调查。

https://www.pianshen.com/article/86001369806/

解决：在这里没有cuda10.2的props，只有11.1的，将cuda的版本调高即可。D:00softwareVS2019MSBuildMicrosoftVCv160BuildCustomizations

(程序抛出的异常) 有错误代码 Runtime Error

https://zhuanlan.zhihu.com/p/360727546
cuda的Runtime API全部都会带有返回值，其类型为cudaError_t；而cuFFT API全部带有cufftResult类型的返回值，curand API则全部带有curandStatus_t类型返回值。

错误代码 cudaSuccess = 0，“no error”

不过要注意的是，__global__函数是异步执行的，如果需要与CPU同步，还需要使用cudaDeviceSynchronize()函数实现同步。所以如果调用完核函数后马上调用cudaGetLastError()，很可能返回cudaSuccess，但核函数运行到某一位置时仍然报错。建议将cudaDeviceSynchronize()函数放在cudaGetLastError()前（但不要用cudaDeviceSynchronize()直接替换掉cudaGetLastError()，有某些异常如cudaErrorInvalidConfiguration并不会在cudaDeviceSynchronize()中报错，而是在cudaGetLastError()中被返回）。

cudaErrorInvalidValue = 1，“invalid argument”

大概率是指针问题。请检查报错的函数传递的参数是不是空指针或野指针，是不是错把指向host端内存的指针当作指向device端内存的指针（或相反）传进了API等。此外还有传入API参数时超过了其范围，如不正常值的枚举等。

1. 使用空指针或野指针
2. 搞混host指针与device指针
3. 使用已经释放的指针
4. 使用错误的枚举值

出现这类问题后，cuda仍可继续提供服务

cudaErrorMemoryAllocation = 2，“out of memory”

调用cudaMalloc时超过了堆内存的最大限制，当然cudaHostAlloc、cudaMallocHost等都会返回这一问题，核函数内的malloc也有产生这一问题的可能性。此时要使用cudaDeviceSetLimit增大堆内存上限，或检查是否发生了内存泄漏，并及时Free掉多余的内存。
出现这类问题后，cuda仍可继续提供服务，仅拒绝分配给用户所申请的内存空间而已。

cudaErrorInitializationError = 3，“initialization error”

正如其名，初始化错误。由于cuda使用了lazy context initialization，这一错误在任何API中都可以返回，但只有第一次被调用时会返回。
出现这类问题后，程序必须终止后重启才能重新使用cuda服务，毕竟初始化只会进行一次，失败了就也办法重来了。
出现这类问题，往往是cuda动态或静态库文件被误删造成的，重装cuda几乎一定可以解决。

cudaErrorInvalidConfiguration = 9，“invalid configuration argument”

运行时参数传递得太大了。

cudaErrorInvalidPitchValue = 12，“invalid pitch argument”

这个错误只会出现在cudaMemcpy2D、cudaMemcpy2DFromArray、cudaMemcpy2DToArray、cudaMemcpy3D及其异步形式函数的返回值中（当然也会被cudaGetLastError和cudaPeekAtLastError捕获到）。
Pitch是通过cudaMallocPitch（申请二维数组）、cudaMalloc3D（申请三维数组）时产生的，用于数据对齐，加速寻址访问速度。

调试技巧

事实上，这种做法太过冗长，每次使用cuda函数都要写六行来处理异常，一个函数的代码量甚至会提升2~3倍；同时代码的连贯性也被打破，后期维护起来非常困难。因此我在前面代码量不大的代码中使用了这种方法：

这种方法解决了上述所有问题，但引入了新的问题，如果发生Assert Failed，我们需要获得异常的返回值来进行处理，因此需要修改代码。但如果发生了无法重现（Unable to Reproduce）的错误，在多次后续实验中难以再次触发，白白浪费了潜在的漏洞修复时机。作为折衷，我们可以定义如下的宏和函数：

#define HANDLE_ERROR(err) (HandleError(err, __FILE__, __LINE__))

//使用HANDLE_ERROR(cudaSetDevice(0));

static void HandleError(cudaError_t err, const char *file, int line) {
	if (err != cudaSuccess) {
		fprintf(stderr, "Error %d: "%s" in %s at line %dn", int(err), cudaGetErrorString(err), file, line);
		exit(int(err));
	}
}

内存问题 CUDA细节 随机数生成

host端当然可以用cstdlib中的rand()函数生成随机数，但设备端如何使用这些随机数？每调用一次rand()，就通过cudaMemcpy传递给显存吗？显然不是，这会消耗太多I/O时间；先调用n次，一次性传到device中吗？
主机仅仅传给设备一个信号，使得多个随机数在device端被生成，这样就能避免过多的I/O带来的耗时问题；或者调用一个设备端函数，使得设备端可以根据需要动态生成随机数。