CUDA代码笔记（一）SimpleMultiGPU

绕梁九日

于 2013-09-11 21:49:52 发布

阅读量4.8k

点赞数 2

CC 4.0 BY-SA版权

分类专栏： CUDA 文章标签： CUDA GPU 异构计算性能优化

本文链接：https://blog.youkuaiyun.com/csgxy123/article/details/11579721

本文是CUDA 5.5样例代码simpleMultiGPU的分析，探讨了如何利用多个GPU进行并行计算。内核函数是GPU的核心任务，main函数通过cudaGetDeviceCount获取GPU数量，并让每个GPU执行内核函数计算部分和。使用getLastCudaError检查错误，通过cudaStreamSynchronize确保GPU计算完成后再进行CPU端的和运算。

摘要生成于 C知道，由 DeepSeek-R1 满血版支持，前往体验 >

这一系列的文章是CUDA5.5样例代码的阅读笔记，每篇文章针对某一特定的样例代码。

本篇文章所涉及的项目是simpleMultiGPU。

项目所在的位置：0_Simple/simpleMultiGPU

源程序的分析

首先我们查看该程序的内核函数（kernel function），以了解程序中GPU所完成的核心功能：

__global__ static void reduceKernel(float *d_Result, float *d_Input, int N) {
    const int tid = blockIdx.x * blockDim.x + threadIdx.x;
    const int threadN = gridDim.x * blockDim.x;
    float sum = 0;

    for (int post = tid; pos < N; pos += threadN)
        sum += d_Input[pos];

    d_Result[tid] = sum;
}

可以看出，threadN代表的是当前GPU中运行的总的线程数，每个线程所完成的工作，是统计数组中所有下标index满足index % threadN == tid的元素的累加和。

然后我们定位到main函数，我们首先查看main函数的前数十行：

// Solver config
TGPUplan plan[MAX_GPU_COUNT];

// GPU reduction results
float h_SumGPU[MAX_GPU_COUNT];

float sumGPU;
double sumCPU, diff;

int i, j, gpuBase, GPU_N;

const int