优化网站教程,淘宝网的公司注册名称,免费空间 个人网站 google广告联盟,安卓系统开发工具CUDA code700(cudaErrorIllegalAddress) 报错与排查方法
最近笔者在调试自己写的 CUDA 代码时, 遇到了 code700(cudaErrorIllegalAddress) 的报错, 在此记录一下排查和解决方法.
报错
报错是由 CUDA API 函数执行时产生的, 由 checkCudaErrors() 函数检测出(CUDA 常用错误检…CUDA code700(cudaErrorIllegalAddress) 报错与排查方法
最近笔者在调试自己写的 CUDA 代码时, 遇到了 code700(cudaErrorIllegalAddress) 的报错, 在此记录一下排查和解决方法.
报错
报错是由 CUDA API 函数执行时产生的, 由 checkCudaErrors() 函数检测出(CUDA 常用错误检测实现, 如下所示).
template typename T
void check(T result, char const *const func, const char *const file,int const line) {if (result) {fprintf(stderr, CUDA error at %s:%d code%d(%s) \%s\ \n, file, line,static_castunsigned int(result), cudaGetErrorName(result), func);exit(EXIT_FAILURE);}
}
#define checkCudaErrors(val) check((val), #val, __FILE__, __LINE__)代码运行时报错如下所示, 显示是执行 cudaMemGetInfo() 函数时错误.
huanghynode8:~/CL/src/cuda/build$ ./example
[sample_cuda] start
[sample_kernel] grid_size:1, block_size:512, shm_size:6144
[sample_kernel] finished
CUDA error at /home/huanghy/CL/src/cuda/sample.cu:53 code700(cudaErrorIllegalAddress) cudaMemGetInfo(freeMem, totalMem) 原因
简单查阅资料可知, code700(cudaErrorIllegalAddress) 的报错原因是 “an illegal memory access was encountered”, 即遇到了一个非法的内存访问.
大多数情况下, 该问题产生都与数组越界访问的情况有关, 但值得一提的是, 往往报错的地方并不是问题实际存在的地方, 而由之前的 kernel 代码中的错误访问导致的. 比如, 此处报错是在 API 函数cudaMemGetInfo() 执行时, 也有可能是在自己定义的 kernel 函数执行时, 但可能一直排查当前报错的 kernel 不能解决问题的.
排查
一个很好的排查上述问题, 也是对自己的 CUDA 代码进行内存访问检查的方法是使用 CUDA 的 compute-sanitizer 工具. 该工具功能很多, 其中一个功能就是进行内存检测.
使用如下指令进行内存检查:
compute-sanitizer --launch-timeout0 --toolmemcheck ./example opt.txt 21其中, ./example 为检测的可执行文件. 由于输出可能比较多, 所以这里重定向到文件中. --launch-timeout0 是将等待 kernel 加载的时间设置为无限, 以避免 compute-sanitizer 出现终止的情况, 如下所示. COMPUTE-SANITIZERError: No attachable process found. compute-sanitizer timed-out.Default timeout can be adjusted with --launch-timeout. Awaiting target completion.最终 compute-sanitizer 会输出检测到的内存访问错误, 如下所示: COMPUTE-SANITIZER
[sample_cuda] start
[sample_kernel] grid_size:1, block_size:512, shm_size:6144Invalid __global__ write of size 4 bytesat 0x1190 in sample_kernel(int *, at::GenericPackedTensorAccessorint, (unsigned long)1, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, (unsigned long)1, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, (unsigned long)1, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, (unsigned long)1, at::RestrictPtrTraits, int, curandStateXORWOW *, unsigned int, int, unsigned int)by thread (32,0,0) in block (0,0,0)Address 0x7f40c00275a4 is out of boundsand is 23,461 bytes after the nearest allocation at 0x7f40c001fc00 of size 7,680 bytesSaved host backtrace up to driver entry point at kernel launch timeHost Frame: [0x305c18]in /usr/lib/x86_64-linux-gnu/libcuda.so.1Host Frame: [0x1488c]in /usr/local/cuda-11.8/lib64/libcudart.so.11.0Host Frame:cudaLaunchKernel [0x6c318]in /usr/local/cuda-11.8/lib64/libcudart.so.11.0Host Frame:cudaError cudaLaunchKernelchar(char const*, dim3, dim3, void**, unsigned long, CUstream_st*) [0x1f2f7]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:__device_stub__Z23sample_kernelPiN2at27GenericPackedTensorAccessorIiLm1ENS0_17RestrictPtrTraitsEiEES3_S3_S3_P17curandStateXORWOWjij(int*, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, curandStateXORWOW*, unsigned int, int, unsigned int) [0x1aec2]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:sample_kernel(int*, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, at::GenericPackedTensorAccessorint, 1ul, at::RestrictPtrTraits, int, curandStateXORWOW*, unsigned int, int, unsigned int) [0x1af3a]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:sample_cuda(std::vectorat::Tensor, std::allocatorat::Tensor , std::vectorCSR, std::allocatorCSR , at::Tensor, CSR const, unsigned int, unsigned int, unsigned int, unsigned long long) [0x1a57d]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:sample(std::vectorat::Tensor, std::allocatorat::Tensor , std::vectorCSR, std::allocatorCSR , at::Tensor, CSR const, unsigned int, unsigned int, unsigned int, unsigned long long) [0x18900]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:main [0x87d0]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:__libc_start_main [0x21c87]in /lib/x86_64-linux-gnu/libc.so.6Host Frame:_start [0x804a]in /home/huanghy/CL/src/cuda/build/./example下面有很多重复单不同 threadIdx 和 blockIdx 的报错, 在此省略[sample_kernel] finishedProgram hit cudaErrorLaunchFailure (error 719) due to unspecified launch failure on CUDA API call to cudaMemGetInfo.Saved host backtrace up to driver entry point at errorHost Frame: [0x4545f6]in /usr/lib/x86_64-linux-gnu/libcuda.so.1Host Frame:cudaMemGetInfo [0x533ab]in /usr/local/cuda-11.8/lib64/libcudart.so.11.0Host Frame:print_device_mem() [0x19796]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:sample_cuda(std::vectorat::Tensor, std::allocatorat::Tensor , std::vectorCSR, std::allocatorCSR , at::Tensor, CSR const, unsigned int, unsigned int, unsigned int, unsigned long long) [0x1a5bb]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:sample(std::vectorat::Tensor, std::allocatorat::Tensor , std::vectorCSR, std::allocatorCSR , at::Tensor, CSR const, unsigned int, unsigned int, unsigned int, unsigned long long) [0x18900]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:main [0x87d0]in /home/huanghy/CL/src/cuda/build/./exampleHost Frame:__libc_start_main [0x21c87]in /lib/x86_64-linux-gnu/libc.so.6Host Frame:_start [0x804a]in /home/huanghy/CL/src/cuda/build/./exampleCUDA error at /home/huanghy/CL/src/cuda/sample.cu:53 code719(cudaErrorLaunchFailure) cudaMemGetInfo(freeMem, totalMem) Target application returned an errorERROR SUMMARY: 34 errors在输出中, compute-sanitizer 会指明在具体的哪个 kernel 函数中发生了越界访问, 并指明相关的 threadIdx 和 blockIdx 以及内存地址.
以上述输出为例, 可以看到是在 sample_kernel() 函数中 threadIdx 为 (32,0,0) blockIdx (0,0,0) 处出现了 Address 0x7f40c00275a4 is out of bounds 的越界访问问题.
虽然地址信息很难让我们确定具体越界访问的位置, 但是通过该工具的输出, 可以确定到具体的 kernel 函数, 对于问题排查已经有了很大帮助.
参考
cuda - Unspecified launch failure on Memcpy - Stack OverflowCompute-sanitizer not quite a drop-in replacement of cuda-memcheck - CUDA Developer Tools / Compute Sanitizer - NVIDIA Developer Forums
