【CUDA代码实践02】矩阵加法运算程序
文章目录
- main.cu
- 代码工作流程
- matrixSum.cuh
- matrixSum.cu
- 代码结构说明
- 总体工作流程
近年来，人工智能（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。
- 在主机端（CPU）为三个向量
-
初始化数据：在主机上初始化
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 的内存初始化为 0memset(fpHost_B, 0, stBytesCount); // 将 B 的内存初始化为 0memset(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 初始化为 0cudaMemset(fpDevice_B, 0, stBytesCount); // 将设备内存 B 初始化为 0cudaMemset(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 和计算出的全局索引 idprintf("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 设备// 检查是否成功设置 GPUif (error != cudaSuccess) {printf("cudaSetDevice failed! \n"); // 如果设置失败，打印错误信息并退出exit(-1);}else {printf("cudaSetDevice success! \n"); // 成功设置 GPU}
}