cudaGetErrorString与cudaGetLastError组合运用

cudaGetErrorString与cudaGetLastError组合运用

前言

在利用集成开发环境(如：visual studio)编写CUDA代码时，编译时是分开编译的，CUDA部分用的是nvcc编译器，这样VS就无法反馈我们CUDA代码的错误，有时候造成了不可避免的内存泄露或者程序出错，自己都不知道，只能一个断点一个断点调试。因此，CUDA也提供了一些函数可以捕捉到error，让我们Debug的时候更方便一点。

cudaGetLastError

通过翻阅CUDA手册，我们可知宿主线程维护着一个初始化为cudaSuccess的cudaError_t类型变量，但出现错误时，该变量都会被错误代码替换。在调用cudaGetLastError()时，就会返回此变量，并将其置为cudaSucess。即返回的是cudaGetLastError()上面代码的错误。
也就是cudaGetLastError可以返回一个变量，该变量存储的是错误代码的信息

cudaGetErrorString

__host__​__device__​const char* 	cudaGetErrorString ( cudaError_t error )

该函数的参数为一个存储错误代码的cudaError_t变量，返回的是该变量对应的错误信息，以字符串的形式接收。

使用方法

1、用于有不需要同步的runtime Api

cudaMalloc(...)
printf("cudaMalloc function : %s\n",cudaGetErrorString(cudaGetLastError()));

这样就可查看cudaMalloc函数是否运行成功。
其实每个runtime API函数都有一个cudaError_t的返回值，我们可以直接将其与cudaSuccess比较即可，可以自己实现一个简单的检查错误。有一些runtime API会自己报错，无需我们检查。

2、用于查找核函数是否有错误(主要功能，CUDA手册介绍的)

核函数不会也不能返回任何错误代码，所以我们必须在核函数启动之后调用cudaGetLastError来检索核函数是否启动成功。同时，我们也还需在核函数启动之前也调用cudaGetLastError来确保启动前的错误变量为cudaSuccess，排除核函数以外的错误信息。
** 由于核函数的启动是异步的，所以需要在启动之后和调用cudaGetLastError之间进行同步 **，否则将无法得到正确的信息。
例如：

printf("befor kernel function : %s\n",cudaGetErrorString(cudaGetLastError()));
kernelfunction<< <...,...> >>();
cudaDeviceSynchronize();//这个同步不能省略
printf("after kernel function : %s\n",cudaGetErrorString(cudaGetLastError()));

这样我们就可以判断核函数启动是否成功了。

3、用于查找其它库的函数是否有错误

当运用到其它库时，也可以使用此方法知道该函数是否运行成功，使用方法和runtime API一样。但是无法准确知道到底时什么错误原因。最好还是用该库对应的错误变量，然后比对该类型的枚举代表的是什么意思。
例如：cublas库的函数返回的变量为cublasStatus_t类型，是个枚举类型。

typedef enum{
    CUBLAS_STATUS_SUCCESS         =0,
    CUBLAS_STATUS_NOT_INITIALIZED =1,
    CUBLAS_STATUS_ALLOC_FAILED    =3,
    CUBLAS_STATUS_INVALID_VALUE   =7,
    CUBLAS_STATUS_ARCH_MISMATCH   =8,
    CUBLAS_STATUS_MAPPING_ERROR   =11,
    CUBLAS_STATUS_EXECUTION_FAILED=13,
    CUBLAS_STATUS_INTERNAL_ERROR  =14,
    CUBLAS_STATUS_NOT_SUPPORTED   =15,
    CUBLAS_STATUS_LICENSE_ERROR   =16
} cublasStatus_t;

通过这个来比对，具体的错误是什么。然后想知道字面上是代表是什么含义的话，可以翻阅CUDA手册。

最后（优化）

如果在大项目里面使用该方法Debug，无疑是很冗长的，所以需要我们对其进行优化。我们可以将其写成宏定义，抽象出其中重复的部分，写成一个宏，然后再封装成一个比较简洁的宏，这样代码就清晰许多。
我们知道这些调试代码会对程序的运行速度大打折扣，所以当发布release版本时，可以将其去掉。这个操作可以用到预编译去完成，我们只需要在Debug模式下调试即可。