代碼分析與調(diào)試方法

CUDA代碼的運行時可能會遇到內(nèi)存溢出、內(nèi)存非法訪問、核函數(shù)出錯等一系列問題，需要進行Debug調(diào)試。在CUDA代碼的分析與調(diào)試過程中，通?？梢圆捎靡韵聨追N方法。

使用printf打印變量信息

在CUDA程序中，可以使用printf()函數(shù)打印變量信息。但是需要注意，由于CUDA是并行計算架構(gòu)，因此在使用printf()函數(shù)時需要謹(jǐn)慎，避免對性能造成過大的影響。

使用CUDA的錯誤檢查功能

在CUDA中，可以使用錯誤處理函數(shù)檢查CUDA程序哪一個環(huán)節(jié)出現(xiàn)錯誤。在每次CUDA函數(shù)調(diào)用后，使用該函數(shù)進行檢查，并打印錯誤信息以幫助診斷問題。另外，也可以使用cuda-memcheck工具來檢查內(nèi)存錯誤。

CUDA編程中錯誤處理相關(guān)的函數(shù)一共有4個：

// cudaGetErrorName函數(shù)接受一個錯誤碼，返回錯誤名稱
__host__ __device__ const char* cudaGetErrorName(cudaError_t error)
// cudaGetErrorString函數(shù)接受一個錯誤碼，返回錯誤描述
__host__ __device__ const char* cudaGetErrorString(cudaError_t error)
// cudaGetLastError函數(shù)返回上一次CUDA runtime調(diào)用時的錯誤碼，并將CUDA錯誤碼置為cudaSuccess
__host__ __device__ cudaError_t cudaGetLastError(void)
// cudaPeekAtLastError函數(shù)與上面功能一樣，但不會重置CUDA的錯誤信息
__host__ __device__ cudaError_t cudaPeekAtLastError(void)

使用CUDA-GDB進行調(diào)試

CUDA-GDB是一個基于命令行的調(diào)試工具，可用于調(diào)試CUDA C/C++應(yīng)用程序。使用CUDA-GDB可以設(shè)置斷點、單步調(diào)試、查看變量值等。需要在編譯時使用-g、-G選項來生成可調(diào)試版本。CUDA-GDB使用方法如下：

# 編譯程序的時候我們使用nvcc編譯器，將其編譯成可調(diào)式版本；-g 表示將CPU代碼(host)編譯成可調(diào)式版本，-G表示將GPU代碼(kernel)編譯成可調(diào)式版本。
nvcc -g -G XXX.cu -o XXX
# 進入調(diào)試器
cuda-gdb XXX
# run執(zhí)行
run

使用Nsight進行調(diào)試

Nsight 開發(fā)工具套件提供深入的跟蹤、調(diào)試、評測和其他分析，以優(yōu)化跨 NVIDIA GPU 和 CPU 的復(fù)雜計算應(yīng)用程序，包括 x86、ARM 和 Power 體系結(jié)構(gòu)。一般在裝CUDA時，在圖形界面已安裝好Nsight。用終端打開Nsight編譯器，它可以直接對寫好程序進行編譯和Debug。關(guān)于如何使用Nsight調(diào)試程序，可以參考鏈接。

使用nvprof / nvvp工具

NVIDIA nvprof / nvvp工具是英偉達N卡GPU編程中用于觀察的利器。全稱是NVIDIA Visual Profiler，是由2008年起開始支持的性能分析器。它具有交互性好，利于使用的優(yōu)點，并且可用于分析CUDA應(yīng)用程序的性能瓶頸和優(yōu)化方案，其在安裝好CUDA toolkit后方可使用。

在CUDA程序調(diào)試過程中，Visual Profiler提供每個CUDA函數(shù)調(diào)用的時間分析，它還能給出如何調(diào)用內(nèi)核函數(shù)以及存儲器的使用情況等，有助于定位瓶頸可能出現(xiàn)的位置，并詳細(xì)解釋如何調(diào)用內(nèi)核等。

記錄運行日志時使用命令nvprof，可視化顯示日志時使用命令nvvp。如果只想對某一段代碼進行分析，在目標(biāo)代碼段前后加上 cudaProfilerStart() 和 cudaProfilerStop()。

# 使用 nvprof 在命令行查看分析結(jié)果
nvprof -o out.nvvp XXX
# 對于生成的 out.nvvp 文件，需要在使用 nvvp 進行查看。

以上是一些常用的CUDA代碼調(diào)試的方法，希望對您有所幫助，歡迎在評論區(qū)留言。

參考文獻

https://zhuanlan.zhihu.com/p/559682306