一、內存訪問優化
- 合并內存訪問：確保相鄰線程訪問連續內存地址（全局內存對齊訪問）。
- 優先使用共享內存（Shared Memory）減少全局內存訪問。
- 避免共享內存的Bank Conflict（例如，使用padding或調整訪問模式）。
- 利用常量內存（Constant Memory）加速只讀數據訪問。
- 使用紋理內存（Texture Memory）或表面內存（Surface Memory）優化隨機訪問。
- 減少全局內存的原子操作（atomic operations）競爭。
- 使用向量化加載（如
float4代替4次float加載）。 - 預取數據到共享內存或寄存器（減少延遲）。
- 避免結構體填充（Struct Padding），手動對齊內存。
- 利用L1/L2緩存優化局部性訪問。
- 使用
__restrict__關鍵字消除指針別名。 - 避免全局內存的未對齊訪問（如地址非32/64字節對齊）。
- 利用只讀緩存（Read-Only Cache）通過
__ldg()指令。 - 合并內存事務寬度（32/64/128字節對齊）。
- 減少內存訪問冗余（如多次讀取同一數據時緩存到寄存器）。
二、執行配置與并行策略
- 合理設置Block和Grid尺寸（典型BlockSize為128/256/512）。
- 最大化活躍線程束（Occupancy）數量（使用CUDA Occupancy Calculator）。
- 避免Block大小導致寄存器溢出（Register Spilling）。
- 使用動態并行（Dynamic Parallelism）時控制子內核粒度。
- 避免過度細分Grid（避免大量小Block導致調度開銷）。
- 使用CUDA Stream實現異步并發執行。
- 優先使用線程束內同步（
__syncwarp()代替__syncthreads()）。 - 避免線程束分化（Warp Divergence）：分支條件盡量在Warp內統一。
- 利用線程束洗牌指令（Warp Shuffle）減少共享內存依賴。
- 使用協作組（Cooperative Groups）優化復雜同步模式。
三、指令與計算優化
- 使用快速數學函數（如
__expf()代替expf()）。 - 避免雙精度計算（除非必需），優先單精度（FP32）或半精度（FP16）。
- 利用Tensor Core加速矩陣運算（FP16/FP32混合精度）。
- 使用內聯函數（
__forceinline__）減少函數調用開銷。 - 避免整數除法和模運算（用位運算或乘法代替）。
- 使用
__ldg()指令優化只讀全局內存訪問。 - 利用
#pragma unroll手動展開循環。 - 避免不必要的類型轉換（如
int與float頻繁轉換）。 - 使用融合乘加（FMA）指令優化計算（
a*b + c）。 - 減少條件分支（使用查表法或預測執行）。
四、寄存器與資源管理
- 限制每個線程的寄存器使用量（避免寄存器溢出）。
- 使用局部變量替代重復計算的中間結果。
- 避免過大的內核參數（通過常量內存或全局內存傳遞）。
- 減少共享內存的靜態分配量（動態共享內存更靈活）。
- 優化線程的局部內存（Local Memory）使用（避免數組過大的棧分配）。
五、通信與同步優化
- 減少
__syncthreads()的使用次數。 - 使用原子操作的輕量級替代（如線程束內投票操作）。
- 優先使用塊內通信（Shared Memory）而非全局內存。
- 避免全局同步（如
cudaDeviceSynchronize()）。 - 使用異步內存復制（
cudaMemcpyAsync）與流重疊計算。
六、工具與調試
- 使用Nsight Compute分析內核性能瓶頸。
- 通過
nvprof或Nsight Systems分析時間線。 - 啟用編譯器優化選項（如
-O3、--use_fast_math）。 - 使用
#pragma unroll提示編譯器展開循環。 - 檢查PTX/SASS代碼確認指令級優化。
- 使用
assert()驗證內存訪問合法性（避免非法訪問導致性能下降）。
七、架構特性適配
- 利用Ampere架構的異步拷貝（Async Copy）特性。
- 為Hopper架構優化DPX指令加速動態規劃。
- 針對Volta+架構優化獨立線程調度（Independent Thread Scheduling）。
- 使用CUDA 11+的集群內存（Cluster Memory）特性。
- 適配不同GPU的Shared Memory/L1 Cache比例（如調整
cudaFuncSetCacheConfig）。
八、數值計算優化
- 使用快速近似函數（如
__saturate()代替手動截斷）。 - 避免非規格化數（Denormals）計算（設置FTZ/DAZ標志）。
- 混合精度訓練時使用
__half2加速半精度計算。 - 利用CUDA數學庫（如CUBLAS、CUTLASS）的優化實現。
九、其他關鍵細節
- 避免主機-設備頻繁通信（減少
cudaMemcpy調用）。 - 使用Zero-Copy內存避免顯式拷貝（Pinned Memory）。
- 內核啟動參數盡量通過常量或寄存器傳遞。
- 減少內核啟動次數（合并多個操作為一個內核）。
- 使用模板元編程（Template Metaprogramming）減少運行時分支。
- 優化全局內存的訪問模式（避免跨步訪問）。
- 利用CUDA Graph捕獲異步操作序列。
- 使用
__builtin_assume_aligned提示編譯器內存對齊。 - 避免線程塊內的資源競爭（如共享內存的讀寫沖突）。
- 利用
__launch_bounds__指定內核資源限制。
十、高級技巧
- 使用PTX內聯匯編優化關鍵路徑。
- 實現雙緩沖（Double Buffering）隱藏內存傳輸延遲。
- 利用共享內存實現高效的歸約（Reduction）操作。
- 使用Warp-level原語（如Warp Reduce/Scan）。
- 優化稀疏數據訪問（如使用壓縮格式）。
- 實現核函數融合（Kernel Fusion）減少中間結果存儲。
- 使用持久化線程（Persistent Threads）處理動態負載。
- 針對數據局部性優化數據布局（如結構體數組轉數組結構體）。
- 利用CUDA的MPS（Multi-Process Service）多進程共享GPU。
- 使用NVTX標記代碼段以輔助性能分析。
十一、常見陷阱
- 誤用共享內存導致Bank Conflict。
- 未初始化共享內存或寄存器變量。
- 線程同步不足導致競態條件。
- 過度使用全局內存原子操作。
- 忽略編譯器警告（如未使用的變量）。
- 錯誤的內存對齊導致性能下降。
- 未優化控制流（如多層嵌套循環）。
- 忽略線程束分化對性能的影響。
- 寄存器溢出導致Local Memory使用。
- 未適配目標GPU的架構限制（如最大線程數）。
十二、其他
- Power of Two: Choosing block sizes that are powers of two (e.g., 128, 256, 512) often leads to better performance due to alignment and coalesced memory accesses.
- Divisibility: Ensure that the total number of threads is divisible by the warp size (32) to avoid underutilization.
