OpenCV-CUDA 图像处理

CUDA编程

GPU并行计算

架构本质与定位

1. GPU的从属角色：GPU并非独立计算平台，而是CPU的协处理器，需与CPU协同工作，因此GPU并行计算的本质是基于CPU+GPU的异构计算架构。
2. 核心组成与连接：CPU（主机端/host）与GPU（设备端/device）通过PCIe总线连接，共同构成异构计算体系。

CPU与GPU的核心差异

二者在核心数量、适用任务、线程特性上存在显著区别，具体对比如下：

异构架构的优势互补逻辑

CPU与GPU通过分工协作发挥最大功效，具体分工为：

• CPU：负责处理逻辑复杂的串行程序，承担任务调度、逻辑判断等核心控制工作。
• GPU：重点处理数据密集型的并行计算程序，利用多核心优势高效完成大规模重复简单运算。

CUDA编程的基础知识和核心概念

1. 内核 (Kernel) 单指令多线程

内核是在GPU上执行的函数，是CUDA并行计算的核心。它由__global__关键字声明，只能从主机(CPU)调用，但在设备(GPU)上执行。

特点：

• 内核启动时会产生大量并行线程，所有线程执行相同的内核函数（单指令多线程SIMT模型）
• 线程通过索引区分自己需要处理的数据
• 内核启动语法：kernel_name<<<gridDim, blockDim, sharedMem, stream>>>(args)
  • gridDim：网格维度（线程块数量）
  • blockDim：线程块维度（每个线程块中的线程数量）
  • sharedMem：每个线程块分配的共享内存大小（可选）
  • stream：指定流（可选，用于异步执行）

示例：

// 声明一个内核函数
__global__ void myKernel(int *data) {
   
   
    // 内核函数体
}

// 启动内核，1024个线程块，每个线程块256个线程
myKernel<<<1024, 256>>>(d_data);

2. 线程 (Thread)

线程是GPU上的基本执行单位，是并行计算的最小单元。每个线程独立执行内核函数，拥有自己的私有内存（寄存器、局部内存）。

特点：

• 每个线程有唯一的标识符，用于确定自己需要处理的数据
• 线程执行时通过内置变量获取自身索引：
  • threadIdx.x：线程在块内的x维度索引
  • threadIdx.y：线程在块内的y维度索引（用于2D线程块）
  • threadIdx.z：线程在块内的z维度索引（用于3D线程块）
• 线程的执行顺序不确定，不能假设执行顺序

线程索引计算示例：

__global__ void kernel(int *array, int n) {
   
   
    // 计算1D全局索引
    int globalIndex = blockIdx.x * blockDim.x + threadIdx.x;
    
    // 确保不越界访问
    if (globalIndex < n) {
   
   
        array[globalIndex] = ...; // 处理数据
    }
}

3. 线程块 (Block)

线程块是一组线程的集合（最多1024个线程），块内线程可以：

• 通过共享内存(__shared__)交换数据
• 通过同步函数(__syncthreads())协调执行

特点：

• 线程块具有三维索引：blockIdx.x, blockIdx.y, blockIdx.z
• 块内线程共享__shared__内存（比全局内存快得多）
• 线程块大小通常选择2的幂（如128、256、512、1024）以优化性能
• 同一线程块的线程必须在同一流式多处理器(SM)上执行
• __syncthreads()用于块内线程同步，确保所有线程都完成某阶段后再继续

线程块示例：

__global__ void sharedMemoryKernel(float *input, float *output, int n) {
   
   
    // 声明共享内存（每个线程块有一份）
    __shared__ float s_data[256];
    
    // 计算索引
    int tid = threadIdx.x;
    int globalIdx = blockIdx.x * blockDim.x + tid;
    
    // 加载数据到共享内存
    s_data[tid] = input[globalIdx];
    
    // 同步：确保所有线程都加载完数据
    __syncthreads();
    
    // 处理共享内存中的数据（比全局内存快）
    output[globalIdx] = s_data[tid] * 2.0f;
}

4. 网格 (Grid)

网格是一组线程块的集合，一个内核启动对应一个网格。网格中的线程块可以在GPU的多个流式多处理器(SM)上并行执行。

特点：

• 网格可以是1D、2D或3D结构，便于处理多维数据（如图像、矩阵）
• 线程块之间不能直接通信或同步，必须通过全局内存间接通信
• 网格中的线程块数量理论上没有上限（受GPU内存限制）
• 全局内存是网格中所有线程共享的内存空间

2D网格和2D线程块示例（适合图像处理）：

__global__ void imageProcessingKernel(unsigned char *input, unsigned char *output, int width, int height) {