跳到主要内容

CUDA 知识速查表

本页面汇总了 CUDA 编程中最常用的语法和知识点，方便快速查阅。

核函数修饰符​

__global__ void kernel() { }    // 核函数，Host 调用，Device 执行
__device__ void deviceFunc() { } // 设备函数，Device 调用，Device 执行
__host__ void hostFunc() { }     // 主机函数，Host 调用，Host 执行
__host__ __device__ void both() { } // 同时在 Host 和 Device 可用

执行配置​

// 一维
kernel<<<gridSize, blockSize>>>(args);
kernel<<<gridSize, blockSize, sharedMemSize>>>(args);
kernel<<<gridSize, blockSize, sharedMemSize, stream>>>(args);

// 二维
dim3 blockSize(16, 16);
dim3 gridSize((width + 15) / 16, (height + 15) / 16);
kernel<<<gridSize, blockSize>>>(args);

// 三维
dim3 blockSize(8, 8, 8);
dim3 gridSize((nx + 7) / 8, (ny + 7) / 8, (nz + 7) / 8);
kernel<<<gridSize, blockSize>>>(args);

内置变量​

变量类型说明
threadIdxuint3线程在 Block 中的索引
blockIdxuint3Block 在 Grid 中的索引
blockDimdim3Block 的维度
gridDimdim3Grid 的维度
warpSizeintWarp 大小（32）
// 计算全局索引
int idx = blockIdx.x * blockDim.x + threadIdx.x;

// 二维索引
int x = blockIdx.x * blockDim.x + threadIdx.x;
int y = blockIdx.y * blockDim.y + threadIdx.y;
int idx = y * width + x;

内存管理​

分配与释放​

cudaMalloc(&d_ptr, size);
cudaFree(d_ptr);
cudaMallocHost(&h_ptr, size);  // 页锁定内存
cudaFreeHost(h_ptr);
cudaMallocManaged(&ptr, size); // 统一内存

内存拷贝​

cudaMemcpy(dst, src, size, cudaMemcpyHostToDevice);
cudaMemcpy(dst, src, size, cudaMemcpyDeviceToHost);
cudaMemcpyAsync(dst, src, size, kind, stream);
cudaMemcpy2D(dst, dpitch, src, spitch, width, height, kind);

内存初始化​

cudaMemset(d_ptr, value, size);
cudaMemsetAsync(d_ptr, value, size, stream);

内存类型​

类型修饰符作用域速度
寄存器自动变量线程最快
共享内存__shared__Block很快
全局内存cudaMallocGrid慢
常量内存__constant__Grid中等
纹理内存texture<>Grid中等
__shared__ float sData[256];
extern __shared__ float dynamicShared[];
__constant__ float constData[256];

同步机制​

__syncthreads();           // Block 内线程同步
__threadfence_block();     // Block 内内存屏障
__threadfence();           // Device 内内存屏障
__threadfence_system();    // 系统级内存屏障
__syncwarp(mask);          // Warp 同步

原子操作​

atomicAdd(addr, value);
atomicSub(addr, value);
atomicExch(addr, value);
atomicMin(addr, value);
atomicMax(addr, value);
atomicInc(addr, value);
atomicDec(addr, value);
atomicCAS(addr, compare, value);
atomicAnd(addr, value);
atomicOr(addr, value);
atomicXor(addr, value);

Warp 级原语​

// 数据交换
__shfl_sync(mask, var, srcLane);
__shfl_up_sync(mask, var, delta);
__shfl_down_sync(mask, var, delta);
__shfl_xor_sync(mask, var, laneMask);

// 投票
__all_sync(mask, predicate);
__any_sync(mask, predicate);
__ballot_sync(mask, predicate);

流与事件​

// 流管理
cudaStream_t stream;
cudaStreamCreate(&stream);
cudaStreamDestroy(stream);
cudaStreamSynchronize(stream);

// 事件管理
cudaEvent_t event;
cudaEventCreate(&event);
cudaEventRecord(event, stream);
cudaEventSynchronize(event);
cudaEventElapsedTime(&ms, start, stop);
cudaEventDestroy(event);

错误处理​

cudaError_t err = cudaGetLastError();
const char* msg = cudaGetErrorString(err);

// 检查宏
#define CUDA_CHECK(call) \
    do { \
        cudaError_t err = call; \
        if (err != cudaSuccess) { \
            fprintf(stderr, "CUDA error: %s\n", cudaGetErrorString(err)); \
            exit(EXIT_FAILURE); \
        } \
    } while(0)

设备属性​

cudaDeviceProp prop;
cudaGetDeviceProperties(&prop, device);

int deviceCount;
cudaGetDeviceCount(&deviceCount);

cudaSetDevice(device);
cudaGetDevice(&device);

常用设备属性​

属性说明
prop.name设备名称
prop.maxThreadsPerBlock每 Block 最大线程数
prop.maxThreadsPerMultiProcessor每 SM 最大线程数
prop.multiProcessorCountSM 数量
prop.totalGlobalMem全局内存大小
prop.sharedMemPerBlock每 Block 共享内存大小
prop.regsPerBlock每 Block 寄存器数量
prop.warpSizeWarp 大小
prop.maxBlocksPerMultiProcessor每 SM 最大 Block 数

性能优化要点​

内存访问​

  1. 合并访问：确保 Warp 内线程访问连续地址
  2. 共享内存：缓存频繁访问的数据
  3. 避免 Bank 冲突：使用填充或调整访问模式
  4. 使用只读缓存：__ldg() 或纹理内存

计算优化​

  1. 减少分支分化：避免 Warp 内分支
  2. 循环展开：#pragma unroll
  3. 使用快速数学函数：__sinf、__expf 等
  4. 使用 __restrict__：告诉编译器指针不重叠

占用率优化​

  1. 选择合适的 Block 大小：32 的倍数
  2. 控制寄存器使用：__launch_bounds__
  3. 平衡共享内存使用

常用编译选项​

nvcc -o program program.cu           # 基本编译
nvcc -g -G program.cu                # 调试模式
nvcc -O3 program.cu                  # 优化模式
nvcc -arch=sm_80 program.cu          # 指定架构
nvcc -use_fast_math program.cu       # 快速数学
nvcc -rdc=true program.cu            # 动态并行
nvcc -Xptxas -v program.cu           # 查看资源使用

常见问题排查​

问题可能原因解决方案
核函数不执行配置错误检查 gridSize、blockSize
结果不正确边界未检查添加 if (idx < n)
性能低内存访问不合并优化访问模式
占用率低寄存器过多减少寄存器使用
死锁同步不当检查 __syncthreads()

参考链接​