nvcc 是 NVIDIA CUDA 生態的核心編譯器，負責將 CUDA C/C++ 代碼（混合了主機代碼和設備代碼）編譯為可在 CPU 和 GPU 上運行的二進制文件。它不僅是一個簡單的編譯器，更是一個“編譯驅動程序”，協調多個工具鏈（如主機編譯器、CUDA 設備編譯器、匯編器、鏈接器）完成整個編譯流程。

一、nvcc 核心功能與編譯流程

1. 核心功能

• 分離并處理 主機代碼（CPU 執行，如 main 函數）和 設備代碼（GPU 執行，如 __global__ 函數）。
• 將設備代碼編譯為 GPU 可執行的二進制指令（SASS）或中間代碼（PTX）。
• 協調主機編譯器（如 gcc、cl.exe）處理主機代碼，并鏈接 CPU/GPU 代碼生成最終可執行文件。

2. 編譯流程（簡化版）

源代碼（.cu）↓ 預處理（合并頭文件、展開宏等）
預處理文件（.i）↓ 前端編譯（分離主機/設備代碼）├─ 主機代碼 → 交給主機編譯器（如 gcc）編譯為 CPU 目標文件└─ 設備代碼 → 編譯為 PTX 中間代碼（.ptx）↓ 后端編譯（PTX → GPU 二進制指令 SASS）設備目標文件（.o）↓ 鏈接（合并主機/設備目標文件 + 鏈接 CUDA 庫）
最終可執行文件

二、基礎用法與核心選項

1. 基礎編譯命令

# 直接編譯 .cu 文件為可執行文件
nvcc source.cu -o output# 分步編譯（先編譯為目標文件，再鏈接）
nvcc -c source.cu -o source.o  # 生成目標文件
nvcc source.o -o output        # 鏈接生成可執行文件

2. 架構相關選項（最核心！）

CUDA 代碼需針對特定 GPU 架構優化，核心選項如下：

選項	作用	示例
-arch=sm_xx	指定目標 GPU 的計算能力（生成該架構的 PTX 和 SASS）	-arch=sm_75（圖靈架構）
-code=sm_xx	僅生成指定架構的 SASS 代碼（需與 -arch 配合）	-arch=compute_70 -code=sm_70,sm_75
-gencode=...	生成多種架構的代碼（更靈活的多架構支持）	-gencode arch=compute_70,code=sm_70
--list-gpu-arch	查看當前 CUDA 版本支持的所有架構	nvcc --list-gpu-arch

示例：編譯支持多架構的程序
（同時支持圖靈 sm_75 和安培 sm_86）：

nvcc -gencode arch=compute_75,code=sm_75 \-gencode arch=compute_86,code=sm_86 \source.cu -o output

注意：compute_xx 表示虛擬架構（用于生成 PTX），sm_xx 表示實際架構（用于生成 SASS）。

3. 預處理選項

控制預處理階段（類似 gcc 的 -E、-D 等）：

選項	作用	示例
-E	僅執行預處理，輸出預處理后的代碼	nvcc -E source.cu -o source.i
-D<宏名>	定義宏（編譯時生效）	-DDEBUG（啟用調試宏）
-I<路徑>	添加頭文件搜索路徑	-I./include
-U<宏名>	取消已定義的宏	-UDEBUG

4. 編譯與優化選項

控制編譯階段的優化、調試等行為：

選項	作用	示例
-O0~-O3	優化級別（-O0 無優化，-O3 最高優化，默認 -O1）	-O2
-g	保留調試信息（供 gdb 等工具使用）	-g
-G	生成設備代碼調試信息（會關閉部分優化，僅用于調試）	-G
-lineinfo	保留設備代碼行號信息（供 nvprof 等性能分析工具使用，不影響優化）	-lineinfo
--use_fast_math	啟用快速數學庫（精度略低，速度更快）	--use_fast_math
-Xptxas <選項>	向 PTX 到 SASS 的匯編器傳遞選項（如控制寄存器使用）	-Xptxas -v（輸出匯編詳細信息）

5. 鏈接選項

控制鏈接階段的庫依賴和路徑：

選項	作用	示例
-L<路徑>	添加庫文件搜索路徑	-L/usr/local/cuda/lib64
-l<庫名>	鏈接指定庫（如 CUDA 運行時、cuBLAS 等）	-lcudart（鏈接 CUDA 運行時庫）
-shared	生成動態鏈接庫（.so/.dll）	nvcc -shared -fPIC source.cu -o libxxx.so
-static	靜態鏈接（僅主機代碼，設備代碼無法靜態鏈接）	-static

6. 輸出中間文件選項

生成編譯過程中的中間文件（用于調試或分析）：

選項	作用	示例
-ptx	生成 PTX 中間代碼（GPU 匯編的文本形式）	nvcc -ptx source.cu（輸出 source.ptx）
-cubin	生成特定架構的二進制代碼（.cubin）	nvcc -arch=sm_75 -cubin source.cu
-keep	保留所有中間文件（.i、.ptx、.cubin 等）	nvcc -keep source.cu

三、進階場景與示例

1. 混合主機代碼（C++）與設備代碼（CUDA）

假設項目結構：

project/
├── main.cpp    # 主機代碼（C++）
└── kernel.cu   # 設備代碼（CUDA）

編譯步驟：

# 1. 編譯設備代碼為目標文件（.o）
nvcc -arch=sm_75 -c kernel.cu -o kernel.o# 2. 用 g++ 編譯主機代碼（需包含 CUDA 頭文件）
g++ -c main.cpp -o main.o -I/usr/local/cuda/include# 3. 鏈接（用 nvcc 確保 CUDA 庫正確鏈接）
nvcc kernel.o main.o -o mixed_app

2. 生成動態鏈接庫（.so）

將 CUDA 代碼封裝為可被其他程序調用的動態庫：

# 編譯為動態庫（-fPIC 生成位置無關代碼）
nvcc -arch=sm_75 -shared -fPIC cuda_lib.cu -o libcudax.so# 使用該庫（例如在 C++ 程序中調用）
g++ app.cpp -o app -L. -lcudax -lcudart  # -lcudart 鏈接 CUDA 運行時

3. 交叉編譯（例如在 x86 主機編譯 ARM 目標）

需配合 NVIDIA 交叉編譯工具鏈（如 Jetson 平臺）：

# 假設交叉編譯器為 aarch64-linux-gnu-g++
nvcc -arch=sm_72 \-ccbin aarch64-linux-gnu-g++ \  # 指定主機交叉編譯器-target-cpu-arch aarch64 \      # 目標 CPU 架構source.cu -o output

4. 查看詳細編譯過程（調試編譯問題）

使用 -v 選項輸出編譯全過程（包括調用的工具、參數等）：

nvcc -v source.cu -o output  # 詳細輸出編譯步驟

四、與構建工具結合（Makefile/CMake）

1. 示例 Makefile

NVCC = nvcc
ARCH = -arch=sm_75
OPTFLAGS = -O2 -lineinfo
INCLUDES = -I./include
LIBS = -lcublas -lcufft  # 鏈接 cuBLAS 和 cuFFT 庫SRC = main.cu kernel.cu
OBJ = $(SRC:.cu=.o)
TARGET = cuda_appall: $(TARGET)%.o: %.cu$(NVCC) $(ARCH) $(OPTFLAGS) $(INCLUDES) -c $< -o $@$(TARGET): $(OBJ)$(NVCC) $(OBJ) -o $@ $(LIBS)clean:rm -f $(OBJ) $(TARGET)

2. 示例 CMakeLists.txt

cmake_minimum_required(VERSION 3.18)
project(cuda_example)# 尋找 CUDA 工具包
find_package(CUDA 12.0 REQUIRED)# 設置 GPU 架構
set(CUDA_ARCHITECTURES 75 86)  # 支持 sm_75 和 sm_86# 添加可執行目標
cuda_add_executable(cuda_app main.cu kernel.cu)# 鏈接 CUDA 庫（如 cuBLAS）
target_link_libraries(cuda_app CUDA::cublas)

五、常見問題與調試技巧

1. “nvcc: command not found”

• 原因：CUDA 未安裝或環境變量未配置。
• 解決：添加環境變量（以 CUDA 12.2 為例）：

  export PATH=/usr/local/cuda-12.2/bin:$PATH
  export LD_LIBRARY_PATH=/usr/local/cuda-12.2/lib64:$LD_LIBRARY_PATH
  

2. “unsupported gpu architecture ‘sm_xx’”

• 原因：指定的 sm_xx 不在當前 CUDA 版本支持范圍（如 CUDA 12 不支持 sm_20）。
• 解決：用 nvcc --list-gpu-arch 查看支持的架構，更換為兼容值。

3. 設備函數未定義（“undefined reference to `kernel<<<>>>'”）

• 原因：設備代碼未被正確編譯或鏈接（如用 g++ 直接鏈接 .cu 文件）。
• 解決：確保用 nvcc 編譯 .cu 文件，且鏈接時使用 nvcc 而非主機編譯器。

4. 優化效果不佳

• 技巧：
  • 用 -Xptxas -v 查看寄存器使用情況（寄存器不足會導致性能下降）。
  • 啟用 -O3 和 --use_fast_math（根據精度需求）。
  • 配合 nvprof 或 nsys 分析性能瓶頸：

    nvprof ./output  # 基礎性能分析
    nsys profile ./output  # 更詳細的全系統分析
    

六、總結

nvcc 的核心是“協調主機與設備代碼的編譯流程”，掌握它的關鍵在于：

1. 正確指定 GPU 架構（-arch/-gencode），匹配目標硬件。
2. 合理使用優化選項（-O2/-Xptxas）提升性能。
3. 結合構建工具（Makefile/CMake）管理復雜項目。
4. 利用中間文件（PTX/SASS）和調試選項（-v/-G）解決編譯問題。

通過 nvcc --help 可查看完整選項列表，結合實際項目調試能更快掌握其用法。