一、内存访问优化
- 合并内存访问：确保相邻线程访问连续内存地址（全局内存对齐访问）。
- 优先使用共享内存（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.
