CUDA 内核函数

news2024/9/22 15:49:42

文章目录

一、简介
二、实现代码
三、实现效果
参考资料

一、简介

CUDA为我们提供了可扩展的编程模型：它的核心是三个关键的抽象——线程组的层次结构、共享内存和阻塞同步——它们只是作为一组最小的语言扩展暴露给外部使用人员。这些抽象提供细粒度数据并行性和线程并行性，嵌套在粗粒度数据并行性和任务并行性中。它们引导程序员将问题划分为可以由线程块独立并行解决的粗子问题，并将每个子问题划分为可以由线程块内的所有线程并行协作解决的细子问题。这种分解通过允许线程在解决每个子问题时进行协作来保持语言的表达性，同时支持自动可伸缩性。实际上，每个线程块都可以在GPU内任何可用的多处理器上调度，以任何顺序，并发或顺序，以便编译后的CUDA程序可以在下图所示的任意数量的多处理器上执行，并且只有运行时系统需要知道物理多处理器计数。这种可扩展的编程模型允许GPU架构通过简单地扩展多处理器和内存分区的数量来跨越广泛的市场范围:从高性能爱好者的GeForce GPU和专业的Quadro和Tesla计算产品到各种廉价的主流GeForce GPU(参见支持cuda的GPU以获得所有支持cuda的GPU的列表)。

GPU是围绕一组流多处理器(SMs)构建的(详见硬件实现)。一个多线程程序被分割成相互独立执行的线程块，这样多处理器的GPU会比少处理器的GPU在更短的时间内自动执行程序。

OK，说了这么多也只是说了一些CUDA的设计理念，对于我们真正使用上来讲，还是要聚焦于代码方面。我们首先要明确一个概念CUDA c++扩展了c++语言，它允许我们定义CUDA c++函数，只不过这种函数通过供多个线程使用，称为内核函数，当调用时，它由N个不同的CUDA线程并行执行N次，而不是像常规的c++函数那样只执行一次。

使用方式：内核是使用__global__声明说明符定义的，对于给定的内核调用，执行该内核的CUDA线程数是使用新的<<<…>>>执行配置语法(参见c++语言扩展)。每个执行内核的线程都有一个唯一的线程ID，可以通过内置变量在内核中访问。这里我们举一个简单的例子。

二、实现代码

#include <iostream>
#include <vector>

#include <cuda_runtime.h>
#include <device_launch_parameters.h>

//定义内核函数
__global__ void VecAdd(float* A, float* B, float* C)
{
    int i = threadIdx.x;
    C[i] = A[i] + B[i];
}

int main()
{
    const int N = 1024; // 假设我们处理1024个元素的向量  
    size_t size = N * sizeof(float);

    // 1. 分配主机内存
    std::vector<float> dataA(size), dataB(size), dataC(size);
    for (int i = 0; i < N; ++i) {
        dataA[i] = rand() / (float)RAND_MAX;
        dataB[i] = rand() / (float)RAND_MAX;
    }

    float* h_A = dataA.data();
    float* h_B = dataB.data();
    float* h_C = dataC.data();

    // 2. 分配设备内存  
    float* d_A, * d_B, * d_C;
    cudaMalloc(&d_A, size);
    cudaMalloc(&d_B, size);
    cudaMalloc(&d_C, size);

    // 3. 复制数据到设备  
    cudaMemcpy(d_A, h_A, size, cudaMemcpyHostToDevice);
    cudaMemcpy(d_B, h_B, size, cudaMemcpyHostToDevice);

    // 4. 启动内核  
    VecAdd <<<1, N >> > (d_A, d_B, d_C);      //这里我们只使用一个线程块，N个线程

    // 等待GPU完成工作  
    cudaDeviceSynchronize();

    // 5. 复制结果回主机  
    cudaMemcpy(h_C, d_C, size, cudaMemcpyDeviceToHost);

    // 6. 验证结果（这里只是打印第一个元素作为示例）  
    std::cout << "A[0] = " << h_A[0] << std::endl 
        << "B[0] = " << h_B[0] << std::endl
        << "C[0] = " << h_C[0] << std::endl;

    // 7. 释放资源  
    cudaFree(d_A);
    cudaFree(d_B);
    cudaFree(d_C);
}