【CUDA代码实践02】矩阵加法运算程序

BQ-Code

已于 2024-10-25 10:16:16 修改

阅读量829

点赞数 18

分类专栏： CUDA篇文章标签：矩阵线性代数 c++ cuda

于 2024-10-25 10:01:40 首次发布

本文链接：https://blog.csdn.net/weixin_52677672/article/details/143227072

版权

CUDA篇专栏收录该内容

3 篇文章

订阅专栏

文章目录

近年来，人工智能（AI）技术，尤其是大模型的快速发展，打开了全新的时代大门。对于想要在这个时代迅速成长并提升自身能力的个人而言，学会利用AI辅助学习已经成为一种趋势。不论是国内的文心一言、豆包，还是国外的ChatGPT、Claude，它们都能成为我们编程学习的有力助手。利用AI进行编程学习将大大提升自己的编程学习效率，这里给大家推荐一个我自己在用的集成ChatGPT和Claude的网站（国内可用，站点稳定）：传送门

本章节代码详细注释了矩阵加法如何在GPU中运行，分为三个文件,main.cu ，matrixSum.cuh，matrixSum.cu

在这里插入图片描述

先放上CUDA程序程序基本框架

__global__ void func(args...){
	核函数内容
}
int main(void)
{
	设置GPU设备;
	分配主机和设备内存;
	数据从主机复制到设备;
	调用核函数在设备中进行计算;
	将就计算得到的数据从设备传给主机;
	释放主机与设备内存;
}

main.cu

代码工作流程

设置 GPU：使用 setGPU() 函数选择并设置合适的 GPU 设备。如果系统中没有可用的 CUDA 兼容 GPU，则会打印错误并退出程序。
分配内存：
- 在主机端（CPU）为三个向量 A, B, C 分配内存，并将内存初始化为 0。
- 在设备端（GPU）为向量 A, B, C 分配内存，并初始化为 0。
初始化数据：在主机上初始化 A 和 B 两个向量的值（通过 initiaData 随机生成数据）。
数据传输：将主机端的 A 和 B 向量复制到 GPU 的设备内存中，C 向量也被传输到设备中，以存储计算结果。
并行计算：
- 通过配置 dim3 的线程块大小和网格大小，决定有多少线程执行并行计算。
- 使用 addFromGPU 内核函数在设备上进行向量加法。
- cudaDeviceSynchronize() 确保所有 GPU 线程完成计算后，程序继续执行。
结果传输：将计算结果 C 从设备端复制回主机端，以便在主机上进行打印。
结果输出：打印向量 A, B 和 C 的前 10 个元素，验证结果是否正确。
清理内存：程序结束时，释放主机和设备的内存，重置 GPU 设备，确保资源得到正确释放。

#include "cuda_runtime.h"  // 包含 CUDA 的运行时库头文件
#include <stdio.h>         // 标准输入输出库，用于 printf
#include "./matrixSum.cuh" // 包含自定义的 CUDA 函数声明

int main()
{
    // 设置 GPU 设备
    setGPU();  // 检测并设置可用的 GPU 设备，确保 GPU 准备好执行计算

    // 分配主机内存和设备内存，并初始化
    int iElemCount = 512;  // 设置元素数量，这里假设每个向量有 512 个元素
    size_t stBytesCount = iElemCount * sizeof(float);  // 计算总共需要的字节数（512 个 float）

    // (1) 分配主机内存，并初始化
    float *fpHost_A, *fpHost_B, *fpHost_C;
    fpHost_A = (float *)malloc(stBytesCount);  // 为向量 A 分配主机内存
    fpHost_B = (float *)malloc(stBytesCount);  // 为向量 B 分配主机内存
    fpHost_C = (float *)malloc(stBytesCount);  // 为结果向量 C 分配主机内存
    if (fpHost_A != NULL && fpHost_B != NULL && fpHost_C != NULL) {
        memset(fpHost_A, 0, stBytesCount); // 将 A 的内存初始化为 0
        memset(fpHost_B, 0, stBytesCount); // 将 B 的内存初始化为 0
        memset(fpHost_C, 0, stBytesCount); // 将 C 的内存初始化为 0
    }
    else {
        printf("fail to malloc memory.\n");  // 如果内存分配失败，输出错误并退出程序
        exit(-1);
    }

    // 分配设备内存，并初始化
    float *fpDevice_A, *fpDevice_B, *fpDevice_C;
    cudaMalloc((float **)&fpDevice_A, stBytesCount);  // 为向量 A 分配设备内存
    cudaMalloc((float **)&fpDevice_B, stBytesCount);  // 为向量 B 分配设备内存
    cudaMalloc((float **)&fpDevice_C, stBytesCount);  // 为结果向量 C 分配设备内存
    if (fpDevice_A != NULL && fpDevice_B != NULL && fpDevice_C != NULL) {
        cudaMemset(fpDevice_A, 0, stBytesCount);  // 将设备内存 A 初始化为 0
        cudaMemset(fpDevice_B, 0, stBytesCount);  // 将设备内存 B 初始化为 0
        cudaMemset(fpDevice_C, 0, stBytesCount);  // 将设备内存 C 初始化为 0
    }
    else {
        printf("fail to malloc memory.\n");  // 如果设备内存分配失败，输出错误并退出程序
        free(fpHost_A);
        free(fpHost_B);
        free(fpHost_C);
        exit(-1);
    }

    // 初始化主机数据
    srand(666);  // 设置随机种子，保证每次生成相同的随机数
    initiaData(fpHost_A, iElemCount);  // 初始化主机向量 A 的数据
    initiaData(fpHost_B, iElemCount);  // 初始化主机向量 B 的数据

    // 将主机数据复制到设备
    cudaMemcpy(fpDevice_A, fpHost_A, stBytesCount, cudaMemcpyHostToDevice);  // 将 A 复制到设备
    cudaMemcpy(fpDevice_B, fpHost_B, stBytesCount, cudaMemcpyHostToDevice);  // 将 B 复制到设备
    cudaMemcpy(fpDevice_C, fpHost_C, stBytesCount, cudaMemcpyHostToDevice);  // 将 C 复制到设备

    // 调用核函数在设备中进行计算
    dim3 block(32);  // 定义每个线程块有 32 个线程
    dim3 grid(iElemCount / block.x);  // 计算需要多少个线程块来处理 512 个元素

    // 调用 GPU 核函数，执行向量加法
    addFromGPU<<<grid, block>>>(fpDevice_A, fpDevice_B, fpDevice_C, iElemCount);
    cudaDeviceSynchronize();  // 同步 CPU 和 GPU，确保 GPU 的计算完成

    // 将结果从设备复制回主机
    cudaMemcpy(fpHost_C, fpDevice_C, stBytesCount, cudaMemcpyDeviceToHost);

    // 打印计算结果中的前 10 个元素
    for (int i = 0; i <10; i++) {
        printf("idx=%2d\tmatrix_A:%.2f\tmatrix_B:%.2f\tresult=%.2f\n", i+1, fpHost_A[i], fpHost_B[i], fpHost_C[i]);
    }

    // 释放主机与设备内存
    free(fpHost_A);  // 释放主机内存
    free(fpHost_B);
    free(fpHost_C);
    cudaFree(fpDevice_A);  // 释放设备内存
    cudaFree(fpDevice_B);
    cudaFree(fpDevice_C);
    cudaDeviceReset();  // 重置设备，清理所有资源

    return 0;  // 程序成功结束
}

matrixSum.cuh

#pragma once
#include <stdlib.h>
#include <stdio.h>

extern void setGPU(); // 声明函数，但不定义
extern void addFromCPU(float *A, float *B, float *C, const int N);
extern void initiaData(float *addr, int elemCount);
extern __global__ void addFromGPU(float *A, float *B, float *C, const int N);

matrixSum.cu

代码结构说明

addFromGPU 内核函数：
- 功能：在 GPU 上并行执行向量加法。每个线程处理向量的一个元素。
- 执行逻辑：
  - 通过 blockIdx.x 和 threadIdx.x 计算每个线程在整个网格中的全局索引 id，确保每个线程都对应唯一的数组元素。
  - 使用 if (id < N) 进行边界检查，避免数组越界访问。
  - 使用 printf 打印当前线程的索引信息，用于调试。
  - 每个线程执行向量加法 C[id] = A[id] + B[id]。
initiaData 函数：
- 功能：初始化一个浮点数数组（向量），每个元素是一个随机生成的值，用于测试向量加法。
- 执行逻辑：
  - 使用 rand() 生成随机数，并将其转换为浮点数格式，填充数组 addr。
setGPU 函数：
- 功能：检测系统中可用的 GPU 数量，并设置使用指定的 GPU 设备（默认为设备 0）。
- 执行逻辑：
  - 调用 cudaGetDeviceCount 获取系统中可用的 GPU 数量。
  - 如果没有找到可用的 GPU，程序退出。
  - 调用 cudaSetDevice 设置当前使用的 GPU 设备（设备 0）。
  - 检查 cudaSetDevice 是否成功，如果失败则退出程序。

总体工作流程

通过 setGPU() 检测系统中可用的 GPU 并选择一个 GPU 设备。
调用 initiaData() 函数生成随机数据，初始化向量 A 和 B。
使用 addFromGPU 内核函数在 GPU 上并行执行向量加法，每个线程处理向量的一部分，最终计算出向量 C。

#include "matrixSum.cuh" // 包含自定义头文件，通常用于声明函数和变量，这里可能包含 CUDA 函数的声明

// 内核函数 (kernel function) addFromGPU，负责在 GPU 上进行向量加法
// 参数：
// - float* A: 输入向量 A 的设备指针
// - float* B: 输入向量 B 的设备指针
// - float* C: 输出向量 C 的设备指针
// - const int N: 向量中元素的数量
__global__ void addFromGPU(float *A, float *B, float *C, const int N)
{
    // blockIdx.x 表示当前线程块 (block) 的 x 维索引
    // threadIdx.x 表示当前线程 (thread) 在线程块中的 x 维索引
    const int bid = blockIdx.x;        // 当前线程所在的线程块的索引
    const int tid = threadIdx.x;       // 当前线程在线程块内的索引
    const int id = bid * blockDim.x + tid;  // 计算全局索引 (global index)
    
    // 计算线程在整个线程网格中的唯一ID，用于确定当前线程负责的元素
    // blockDim.x 表示每个线程块中的线程数，这里通过线程块索引 (bid) 和线程索引 (tid) 计算唯一的全局索引

    // 边界检查：确保 id 不会超出向量的大小，避免访问越界的内存
    if (id < N) {
        // 打印当前线程的 block ID、thread ID 和计算出的全局索引 id
        printf("bid = %d\ttid = %d\tid = %d\n", bid, tid, id);
        // 执行向量加法，计算 C[id] = A[id] + B[id]，每个线程负责一个元素的加法
        C[id] = A[id] + B[id];
    }
}

// 初始化数据的函数，将随机生成的数据填充到向量中
// 参数：
// - float* addr: 指向需要初始化的向量的指针
// - int elemCount: 向量中的元素数量
void initiaData(float *addr, int elemCount)
{
    // 循环填充每个元素，生成随机数 (0-255) 的浮点表示，范围在 0 到 25.5 之间
    for (int i = 0; i < elemCount; i++)
    {
        addr[i] = (float)(rand() & 0xFF) / 10.f;  // 取 0 到 255 之间的随机整数，并转换为浮点数
    }
    return;  // 函数结束，返回主程序
}

// 设置 GPU 的函数，用于检测可用 GPU 并选择设备
void setGPU()
{
    int iDeviceCount = 0;  // 存储 GPU 的数量
    cudaError_t error = cudaGetDeviceCount(&iDeviceCount);  // 获取可用的 GPU 设备数量

    // 检查是否存在可用的 CUDA 兼容设备
    if (error != cudaSuccess || iDeviceCount == 0) {
        printf("No CUDA compatible GPU found\n");
        exit(-1);  // 如果没有找到 GPU，程序终止
    }
    else {
        printf("The count of GPUs is %d.\n", iDeviceCount);  // 打印检测到的 GPU 数量
    }

    // 选择设备 0（如果有多个 GPU，这里默认选择第一个 GPU）
    int iDevice = 0;
    error = cudaSetDevice(iDevice);  // 设置当前使用的 GPU 设备

    // 检查是否成功设置 GPU
    if (error != cudaSuccess) {
        printf("cudaSetDevice failed! \n");  // 如果设置失败，打印错误信息并退出
        exit(-1);
    }
    else {
        printf("cudaSetDevice success! \n");  // 成功设置 GPU
    }
}