当前位置： 首页 > news >正文

我的第一个CUDA程序

news 来源：原创 2024/9/22 8:29:02

MatAdd算法

实现两个矩阵对应元素相加

 

#include <stdio.h>
#include <stdlib.h>// 矩阵加法函数
void MatAdd(int height, int width)
{// 在主机内存中为 A、B 和 C 分配内存float* A = (float*)malloc(height * width * sizeof(float));float* B = (float*)malloc(height * width * sizeof(float));float* C = (float*)malloc(height * width * sizeof(float));// 初始化输入矩阵 A 和 Bfor (int i = 0; i < height; i++){for (int j = 0; j < width; j++){// 例如，初始化为 i+j 的和*(A + i * width + j) = i + j;*(B + i * width + j) = i - j;}}// 执行矩阵加法运算，将结果存储在矩阵 C 中for (int i = 0; i < height; i++){for (int j = 0; j < width; j++){*(C + i * width + j) = *(A + i * width + j) + *(B + i * width + j);}}// 输出结果矩阵 Cprintf("Matrix C (result of A + B):\n");for (int i = 0; i < height; i++){for (int j = 0; j < width; j++){printf("%f ", *(C + i * width + j));}printf("\n");}// 释放分配的内存free(A);free(B);free(C);
}int main()
{int height = 3; // 矩阵的高度int width = 3;  // 矩阵的宽度// 执行矩阵加法MatAdd(height, width);return 0;
}

 输出：

Matrix C (result of A + B):
0.000000 0.000000 0.000000 
2.000000 2.000000 2.000000 
4.000000 4.000000 4.000000

MatAdd算法的GPU实现

MatAdd算法的GPU实现

  • CPU端为输入矩阵A和B、输出矩阵C分配空间，并进行初始化
  • CPU端分配设备端内存，并将A和B传输到GPU上
  • 定义数据和线程的映射关系，并确定线程的开启数量和组织方式
  1. 每个线程负责输出矩阵C的一个元素的计算，全局ID为（i，j）的线程计算索引为（i，j）的矩阵元素
  2. 当矩阵规模为width*height时，共开启width * height个线程
  3. 每个block包含256个线程，采用（16,16）的组织形式
  • 编写计算kernel，完成计算任务
  • CPU端将计算结果从Device内存拷贝到Host内存

内存分配 数据传输

开启线程 启动kernel 结果返回

GPU kernel

#include <stdio.h>
#include <cuda_runtime.h>// CUDA 核函数，用于矩阵加法
__global__ void MatAddKernel(float* A, float* B, float* C, int height, int width) {// 获取线程的全局IDint i = blockIdx.x * blockDim.x + threadIdx.x; // 计算全局行索引int j = blockIdx.y * blockDim.y + threadIdx.y; // 计算全局列索引// 确保索引在矩阵范围内if (i < width && j < height) {// 计算当前线程对应的元素索引int index = j * width + i;// 从矩阵 A 和 B 中读取数据float src_data_A = A[index];float src_data_B = B[index];// 执行加法运算float result = src_data_A + src_data_B;// 将结果写入矩阵 CC[index] = result;}
}void MatAdd(int height, int width) {// 在主机内存中分配 A、B 和 Cfloat* A = (float*)malloc(height * width * sizeof(float));float* B = (float*)malloc(height * width * sizeof(float));float* C = (float*)malloc(height * width * sizeof(float));// 初始化输入矩阵 A 和 Bfor (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {A[i * width + j] = i + j; // 简单初始化，A的元素为行索引+列索引B[i * width + j] = i - j; // 简单初始化，B的元素为行索引-列索引}}// 第一步：在设备内存中为矩阵 A、B 和 C 分配内存float* d_A;cudaMalloc(&d_A, height * width * sizeof(float)); // 分配矩阵 A 的设备内存float* d_B;cudaMalloc(&d_B, height * width * sizeof(float)); // 分配矩阵 B 的设备内存float* d_C;cudaMalloc(&d_C, height * width * sizeof(float)); // 分配矩阵 C 的设备内存// 第二步：将矩阵 A 和 B 从主机内存复制到设备内存cudaMemcpy(d_A, A, height * width * sizeof(float), cudaMemcpyHostToDevice); // 复制 AcudaMemcpy(d_B, B, height * width * sizeof(float), cudaMemcpyHostToDevice); // 复制 B// 第三步：调用 CUDA 核函数dim3 threadsPerBlock(16, 16); // 定义每个块中的线程数dim3 numBlocks((width + threadsPerBlock.x - 1) / threadsPerBlock.x, (height + threadsPerBlock.y - 1) / threadsPerBlock.y); // 计算网格中的块数MatAddKernel<<<numBlocks, threadsPerBlock>>>(d_A, d_B, d_C, height, width); // 启动 CUDA 核函数// 第四步：将结果从设备内存复制回主机内存cudaMemcpy(C, d_C, height * width * sizeof(float), cudaMemcpyDeviceToHost);// 输出结果矩阵 Cprintf("Matrix C (result of A + B):\n");for (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {printf("%f ", C[i * width + j]);}printf("\n");}// 释放设备内存cudaFree(d_A);cudaFree(d_B);cudaFree(d_C);// 释放主机内存free(A);free(B);free(C);
}int main() {int height = 3; // 矩阵的高度int width = 3;  // 矩阵的宽度// 调用矩阵加法函数MatAdd(height, width);return 0;
}

输出：

Matrix C (result of A + B):
0.000000 0.000000 0.000000 
2.000000 2.000000 2.000000 
4.000000 4.000000 4.000000

CUDA程序编译

CUDA程序性能测试

 使用 CUDA GPU Timers 实际要循环100次求平均值

#include <stdio.h>
#include <cuda_runtime.h>// CUDA 核函数，用于矩阵加法
__global__ void MatAddKernel(float* A, float* B, float* C, int height, int width) {// 获取线程的全局IDint i = blockIdx.x * blockDim.x + threadIdx.x; // 计算全局行索引int j = blockIdx.y * blockDim.y + threadIdx.y; // 计算全局列索引// 确保索引在矩阵范围内if (i < width && j < height) {// 计算当前线程对应的元素索引int index = j * width + i;// 从矩阵 A 和 B 中读取数据float src_data_A = A[index];float src_data_B = B[index];// 执行加法运算float result = src_data_A + src_data_B;// 将结果写入矩阵 CC[index] = result;}
}void MatAdd(int height, int width) {// 在主机内存中分配 A、B 和 Cfloat* A = (float*)malloc(height * width * sizeof(float));float* B = (float*)malloc(height * width * sizeof(float));float* C = (float*)malloc(height * width * sizeof(float));// 初始化输入矩阵 A 和 Bfor (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {A[i * width + j] = i + j; // 简单初始化，A的元素为行索引+列索引B[i * width + j] = i - j; // 简单初始化，B的元素为行索引-列索引}}// 第一步：在设备内存中为矩阵 A、B 和 C 分配内存float* d_A;cudaMalloc(&d_A, height * width * sizeof(float)); // 分配矩阵 A 的设备内存float* d_B;cudaMalloc(&d_B, height * width * sizeof(float)); // 分配矩阵 B 的设备内存float* d_C;cudaMalloc(&d_C, height * width * sizeof(float)); // 分配矩阵 C 的设备内存// 第二步：将矩阵 A 和 B 从主机内存复制到设备内存cudaMemcpy(d_A, A, height * width * sizeof(float), cudaMemcpyHostToDevice); // 复制 AcudaMemcpy(d_B, B, height * width * sizeof(float), cudaMemcpyHostToDevice); // 复制 B// 第三步：定义事件变量，用于测量核函数执行时间cudaEvent_t start, stop;float time;cudaEventCreate(&start); // 创建开始事件cudaEventCreate(&stop);  // 创建停止事件// 第四步：调用 CUDA 核函数dim3 threadsPerBlock(16, 16); // 定义每个块中的线程数dim3 numBlocks((width + threadsPerBlock.x - 1) / threadsPerBlock.x, (height + threadsPerBlock.y - 1) / threadsPerBlock.y); // 计算网格中的块数cudaEventRecord(start, 0); // 记录开始时间MatAddKernel<<<numBlocks, threadsPerBlock>>>(d_A, d_B, d_C, height, width); // 启动 CUDA 核函数cudaEventRecord(stop, 0);  // 记录结束时间cudaEventSynchronize(stop); // 等待事件完成cudaEventElapsedTime(&time, start, stop); // 计算核函数执行时间printf("Kernel execution time: %f ms\n", time); // 输出核函数执行时间// 销毁事件cudaEventDestroy(start);cudaEventDestroy(stop);// 第五步：将结果从设备内存复制回主机内存cudaMemcpy(C, d_C, height * width * sizeof(float), cudaMemcpyDeviceToHost);// 输出结果矩阵 Cprintf("Matrix C (result of A + B):\n");for (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {printf("%f ", C[i * width + j]);}printf("\n");}// 释放设备内存cudaFree(d_A);cudaFree(d_B);cudaFree(d_C);// 释放主机内存free(A);free(B);free(C);
}int main() {int height = 3; // 矩阵的高度int width = 3;  // 矩阵的宽度// 调用矩阵加法函数MatAdd(height, width);return 0;
}

输出： 

Kernel execution time: 0.086016 ms
Matrix C (result of A + B):
0.000000 0.000000 0.000000 
2.000000 2.000000 2.000000 
4.000000 4.000000 4.000000

 使用 CPU Timers 实际要循环100次求平均值

#include <stdio.h>
#include <stdlib.h>
#include <sys/time.h> // 包含 gettimeofday 的头文件
#include <cuda_runtime.h>// CUDA 核函数，用于矩阵加法
__global__ void MatAddKernel(float* A, float* B, float* C, int height, int width) {// 获取线程的全局IDint i = blockIdx.x * blockDim.x + threadIdx.x; // 计算全局行索引int j = blockIdx.y * blockDim.y + threadIdx.y; // 计算全局列索引// 确保索引在矩阵范围内if (i < width && j < height) {// 计算当前线程对应的元素索引int index = j * width + i;// 从矩阵 A 和 B 中读取数据float src_data_A = A[index];float src_data_B = B[index];// 执行加法运算float result = src_data_A + src_data_B;// 将结果写入矩阵 CC[index] = result;}
}void MatAdd(int height, int width) {// 在主机内存中分配 A、B 和 Cfloat* A = (float*)malloc(height * width * sizeof(float));float* B = (float*)malloc(height * width * sizeof(float));float* C = (float*)malloc(height * width * sizeof(float));// 初始化输入矩阵 A 和 Bfor (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {A[i * width + j] = i + j; // 简单初始化，A的元素为行索引+列索引B[i * width + j] = i - j; // 简单初始化，B的元素为行索引-列索引}}// 第一步：在设备内存中为矩阵 A、B 和 C 分配内存float* d_A;cudaMalloc(&d_A, height * width * sizeof(float)); // 分配矩阵 A 的设备内存float* d_B;cudaMalloc(&d_B, height * width * sizeof(float)); // 分配矩阵 B 的设备内存float* d_C;cudaMalloc(&d_C, height * width * sizeof(float)); // 分配矩阵 C 的设备内存// 第二步：将矩阵 A 和 B 从主机内存复制到设备内存cudaMemcpy(d_A, A, height * width * sizeof(float), cudaMemcpyHostToDevice); // 复制 AcudaMemcpy(d_B, B, height * width * sizeof(float), cudaMemcpyHostToDevice); // 复制 B// 第三步：定义时间测量变量struct timeval start, end;double elapsedTime;// 记录开始时间gettimeofday(&start, NULL);// 启动 CUDA 核函数dim3 threadsPerBlock(16, 16); // 定义每个块中的线程数dim3 numBlocks((width + threadsPerBlock.x - 1) / threadsPerBlock.x, (height + threadsPerBlock.y - 1) / threadsPerBlock.y); // 计算网格中的块数MatAddKernel<<<numBlocks, threadsPerBlock>>>(d_A, d_B, d_C, height, width); // 启动 CUDA 核函数cudaDeviceSynchronize(); // 等待核函数完成// 记录结束时间gettimeofday(&end, NULL);// 计算时间差elapsedTime = (end.tv_sec - start.tv_sec) * 1000.0; // 秒转毫秒elapsedTime += (end.tv_usec - start.tv_usec) / 1000.0; // 微秒转毫秒// 输出核函数执行时间printf("Kernel execution time: %f ms\n", elapsedTime);// 第四步：将结果从设备内存复制回主机内存cudaMemcpy(C, d_C, height * width * sizeof(float), cudaMemcpyDeviceToHost);// 输出结果矩阵 Cprintf("Matrix C (result of A + B):\n");for (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {printf("%f ", C[i * width + j]);}printf("\n");}// 释放设备内存cudaFree(d_A);cudaFree(d_B);cudaFree(d_C);// 释放主机内存free(A);free(B);free(C);
}int main() {int height = 3; // 矩阵的高度int width = 3;  // 矩阵的宽度// 调用矩阵加法函数MatAdd(height, width);return 0;
}

 输出：

Kernel execution time: 0.101000 ms
Matrix C (result of A + B):
0.000000 0.000000 0.000000 
2.000000 2.000000 2.000000 
4.000000 4.000000 4.000000

 MatAdd程序之变一：每个线程处理4个元素

  • CPU端为输入矩阵A和B、输出矩阵C分配空间，并进行初始化
  • CPU端分配设备端内存，并将A和B传输到GPU上
  • 定义数据和线程的映射关系，并确定线程的开启数量和组织方式
  1. 每个线程负责输出矩阵C的四个元素的计算，全局ID为（i，j）的线程计算索引为（i，4*j~4*j+3）的矩阵元素
  2. 当矩阵规模为width*height时，共开启width/4 * height个线程
  3. 每个block包含256个线程，采用（16,16）的组织形式
  • 编写计算kernel，完成计算任务
  • CPU端将计算结果从Device内存拷贝到Host内存

#include <cuda_runtime.h>
#include <iostream>// 自定义向上取整除法函数
int div_up(int a, int b) {return (a + b - 1) / b;
}
// CUDA核函数，执行矩阵加法
__global__ void MatAdd(float *A, float *B, float *C, int height, int width)
{// 获取线程的全局IDint i = blockIdx.x * blockDim.x + threadIdx.x;int j = blockIdx.y * blockDim.y + threadIdx.y;// 检查线程是否在有效的计算范围内// 每个线程处理4个连续的元素，因此检查4*i是否在宽度范围内if ((4 * i) < width && j < height){// 获取线程与数据之间的映射关系int index = j * width + 4 * i;// 从源矩阵读取数据float4 src_data_A = *((float4 *)(A + index));float4 src_data_B = *((float4 *)(B + index));// 执行加法运算// 执行加法运算，逐个对float4的分量进行加法float4 result;result.x = src_data_A.x + src_data_B.x;result.y = src_data_A.y + src_data_B.y;result.z = src_data_A.z + src_data_B.z;result.w = src_data_A.w + src_data_B.w;// 将结果写回结果矩阵*((float4 *)(C + index)) = result;}
}// 矩阵加法函数，处理矩阵的加法运算
void MatAdd(int height, int width)
{float *A, *B, *C;         // 主机内存指针float *d_A, *d_B, *d_C;   // 设备内存指针// 分配主机内存A = (float*)malloc(height * width * sizeof(float));B = (float*)malloc(height * width * sizeof(float));C = (float*)malloc(height * width * sizeof(float));// 初始化矩阵A和B（这里省略具体的初始化代码）// 初始化输入矩阵 A 和 Bfor (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {A[i * width + j] = i + j; // 简单初始化，A的元素为行索引+列索引B[i * width + j] = i - j; // 简单初始化，B的元素为行索引-列索引}}// 分配设备内存cudaMalloc((void**)&d_A, height * width * sizeof(float));cudaMalloc((void**)&d_B, height * width * sizeof(float));cudaMalloc((void**)&d_C, height * width * sizeof(float));// 将数据从主机内存复制到设备内存cudaMemcpy(d_A, A, height * width * sizeof(float), cudaMemcpyHostToDevice);cudaMemcpy(d_B, B, height * width * sizeof(float), cudaMemcpyHostToDevice);// 设置线程块的大小，每个块中有16x16个线程dim3 threadsPerBlock(16, 16);// 计算需要多少个线程块来覆盖整个矩阵区域// 宽度除以4是因为每个线程处理4个元素dim3 numBlocks(div_up(width / 4, threadsPerBlock.x), div_up(height, threadsPerBlock.y));// 调用核函数，执行矩阵加法运算MatAdd<<<numBlocks, threadsPerBlock>>>(d_A, d_B, d_C, height, width);// 将计算结果从设备内存复制到主机内存cudaMemcpy(C, d_C, height * width * sizeof(float), cudaMemcpyDeviceToHost);
// 输出结果矩阵 Cprintf("Matrix C (result of A + B):\n");for (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {printf("%f ", C[i * width + j]);}printf("\n");}// 释放设备内存cudaFree(d_A);cudaFree(d_B);cudaFree(d_C);// 释放主机内存free(A);free(B);free(C);
}int main() {int height = 8; // 示例矩阵的高度int width = 8;  // 示例矩阵的宽度// 调用矩阵加法函数MatAdd(height, width);return 0;
}

 输出：

Matrix C (result of A + B):
0.000000 0.000000 0.000000 0.000000 0.000000 0.000000 0.000000 0.000000 
2.000000 2.000000 2.000000 2.000000 2.000000 2.000000 2.000000 2.000000 
4.000000 4.000000 4.000000 4.000000 4.000000 4.000000 4.000000 4.000000 
6.000000 6.000000 6.000000 6.000000 6.000000 6.000000 6.000000 6.000000 
8.000000 8.000000 8.000000 8.000000 8.000000 8.000000 8.000000 8.000000 
10.000000 10.000000 10.000000 10.000000 10.000000 10.000000 10.000000 10.000000 
12.000000 12.000000 12.000000 12.000000 12.000000 12.000000 12.000000 12.000000 
14.000000 14.000000 14.000000 14.000000 14.000000 14.000000 14.000000 14.000000

MatAdd程序之变二

矩阵A、B、C都为NxN的方阵，A和B为已知矩阵，C[i][j] = A[i][j] + B[j][i]。

 

  • CPU端为输入矩阵A和B、输出矩阵C分配空间，并进行初始化
  • CPU端分配设备端内存，并将A和B传输到GPU上
  • 定义数据和线程的映射关系，并确定线程的开启数量和组织方式
  1. 每个线程负责输出矩阵C的一个元素的计算，全局ID为（i，j）的线程计算索引为（i，j）的矩阵元素
  2. 当矩阵规模为width*height时，共开启width * height个线程
  3. 每个block包含256个线程，采用（16,16）的组织形式
  • 编写计算kernel，完成计算任务
  • CPU端将计算结果从Device内存拷贝到Host内存

#include <cuda_runtime.h>
#include <iostream>#define N 1024 // 矩阵的大小（N x N）// CUDA核函数，执行矩阵加法运算
__global__ void MatAdd(float *A, float *B, float *C, int height, int width)
{// 获取线程的全局IDint i = blockIdx.x * blockDim.x + threadIdx.x;int j = blockIdx.y * blockDim.y + threadIdx.y;// 确保线程在有效的计算范围内if(i < width && j < height){// 获取矩阵元素的索引int index_A = j * width + i;int index_B = i * width + j;int index_C = index_A;// 从矩阵A和B读取数据并计算结果float src_data_A = *(A + index_A);float src_data_B = *(B + index_B);float result = src_data_A + src_data_B;// 将结果写入矩阵C*(C + index_C) = result;}
}// 主函数，处理矩阵的初始化、调用CUDA核函数以及结果输出
void MatAdd(int height, int width)
{// 在主机内存中为A、B和C分配空间float* A = (float*)malloc(height * width * sizeof(float));float* B = (float*)malloc(height * width * sizeof(float));float* C = (float*)malloc(height * width * sizeof(float));// 初始化矩阵A和Bfor (int i = 0; i < height; i++) {for (int j = 0; j < width; j++) {*(A + i * width + j) = static_cast<float>(i + j); // 示例初始化*(B + i * width + j) = static_cast<float>(i - j); // 示例初始化}}// 分配设备内存float *d_A, *d_B, *d_C;cudaMalloc((void**)&d_A, height * width * sizeof(float));cudaMalloc((void**)&d_B, height * width * sizeof(float));cudaMalloc((void**)&d_C, height * width * sizeof(float));// 将数据从主机内存复制到设备内存cudaMemcpy(d_A, A, height * width * sizeof(float), cudaMemcpyHostToDevice);cudaMemcpy(d_B, B, height * width * sizeof(float), cudaMemcpyHostToDevice);// 配置CUDA网格和线程块大小dim3 threadsPerBlock(16, 16);dim3 numBlocks((width + threadsPerBlock.x - 1) / threadsPerBlock.x,(height + threadsPerBlock.y - 1) / threadsPerBlock.y);// 调用CUDA核函数，执行矩阵加法运算MatAdd<<<numBlocks, threadsPerBlock>>>(d_A, d_B, d_C, height, width);// 将计算结果从设备内存复制回主机内存cudaMemcpy(C, d_C, height * width * sizeof(float), cudaMemcpyDeviceToHost);// 输出部分计算结果用于验证std::cout << "C[0][0] = " << C[0] << std::endl;std::cout << "C[N-1][N-1] = " << C[(N-1) * width + (N-1)] << std::endl;// 释放设备内存cudaFree(d_A);cudaFree(d_B);cudaFree(d_C);// 释放主机内存free(A);free(B);free(C);
}int main() {// 调用矩阵加法函数MatAdd(N, N);return 0;
}

输出：

C[0][0] = 0
C[N-1][N-1] = 2046

MatAdd程序之变三

  •  矩阵A、B都为MxN的矩阵，矩阵C为(M/2)*(N/2)的矩阵， A和B为已知矩阵，C[i][j] = A[2*i][2*j] *B[2*i][2*j] +  A[2*i][2*j+1] *B[2*i][2*j+1]   + A[2*i+1][2*j] *B[2*i+1][2*j]  + A[2*i+1][2*j +1] *B[2*i+1][2*j+1] 。 CPU端为输入矩阵A和B、输出矩阵C分配空间，并进行初始化
  • CPU端分配设备端内存，并将A和B传输到GPU上
  • 定义数据和线程的映射关系，并确定线程的开启数量和组织方式
  1. 每个线程负责输出矩阵C的一个元素的计算，全局ID为（i，j）的线程计算索引为（i，j）的矩阵元素
  2. 当矩阵规模为(M/2)*(N/2)时，共开启(M/2)*(N/2)个线程,每个线程对应A和B的四个元素
  3. 每个block包含256个线程，采用（16,16）的组织形式
  • 编写计算kernel，完成计算任务
  • CPU端将计算结果从Device内存拷贝到Host内存

#include <iostream>
#include <cuda_runtime.h>#define BLOCK_SIZE 16__global__ void MatAdd(float *A, float *B, float *C, int height, int width)
{// Get thread IDint i = blockIdx.x * blockDim.x + threadIdx.x;int j = blockIdx.y * blockDim.y + threadIdx.y;if (i < width / 2 && j < height / 2){// Get the mapping between thread and dataint index_A_1 = 2 * j * width + 2 * i;int index_A_2 = 2 * j * width + 2 * i + 1;int index_A_3 = (2 * j + 1) * width + 2 * i;int index_A_4 = (2 * j + 1) * width + 2 * i + 1;int index_B_1 = index_A_1; // Indexes for matrix B are the same as Aint index_B_2 = index_A_2;int index_B_3 = index_A_3;int index_B_4 = index_A_4;int index_C = j * (width / 2) + i;// Read data from source matricesfloat A1 = A[index_A_1];float A2 = A[index_A_2];float A3 = A[index_A_3];float A4 = A[index_A_4];float B1 = B[index_B_1];float B2 = B[index_B_2];float B3 = B[index_B_3];float B4 = B[index_B_4];// Computefloat result = A1 * B1 + A2 * B2 + A3 * B3 + A4 * B4;// Write resultC[index_C] = result;}
}int main()
{int M = 4; // Example dimensions for the matricesint N = 4;int size_A = M * N * sizeof(float);int size_B = M * N * sizeof(float);int size_C = (M / 2) * (N / 2) * sizeof(float);// Allocate host memoryfloat *h_A = (float*)malloc(size_A);float *h_B = (float*)malloc(size_B);float *h_C = (float*)malloc(size_C);// Initialize matrices A and Bfor (int i = 0; i < M * N; ++i){h_A[i] = i + 1; // Example initializationh_B[i] = i + 1;}// Allocate device memoryfloat *d_A, *d_B, *d_C;cudaMalloc(&d_A, size_A);cudaMalloc(&d_B, size_B);cudaMalloc(&d_C, size_C);// Copy data from host to devicecudaMemcpy(d_A, h_A, size_A, cudaMemcpyHostToDevice);cudaMemcpy(d_B, h_B, size_B, cudaMemcpyHostToDevice);// Define block and grid sizedim3 blockDim(BLOCK_SIZE, BLOCK_SIZE);dim3 gridDim((N / 2 + BLOCK_SIZE - 1) / BLOCK_SIZE, (M / 2 + BLOCK_SIZE - 1) / BLOCK_SIZE);// Launch kernelMatAdd<<<gridDim, blockDim>>>(d_A, d_B, d_C, M, N);// Copy result from device to hostcudaMemcpy(h_C, d_C, size_C, cudaMemcpyDeviceToHost);// Print resultfor (int i = 0; i < M / 2; ++i){for (int j = 0; j < N / 2; ++j){std::cout << h_C[i * (N / 2) + j] << " ";}std::cout << std::endl;}// Free device memorycudaFree(d_A);cudaFree(d_B);cudaFree(d_C);// Free host memoryfree(h_A);free(h_B);free(h_C);return 0;
}

 输出：

66 138 
546 746
  • 异构计算整成为当前计算领域的重点方向
  • GPGPU是异构计算的主要形式
  • GPGPU是一款大规模细粒度并行处理器，并行思维是进行GPGPU编程的重要前提
  • NVIDIA是当前GPGPU领域当之无愧的霸主
  • GPGPU编程的重点是定义明确的线程和数据间的映射

相关文章：

  • 北京网站建设多少钱？
  • 辽宁网页制作哪家好_网站建设
  • 高端品牌网站建设_汉中网站制作
  • AI大模型入门基础教程（非常详细），AI大模型入门到精通，收藏这一篇就够了！
  • C# MaterialDesign抽屉式风格
  • 静态代理和动态代理
  • JavaEE从入门到起飞(八) ~ Git
  • 今日（2024 年 8 月 15 日）科技新闻
  • RESTful API设计指南：构建高效、可扩展的Web服务
  • 仿RabbitMq实现简易消息队列正式篇（连接篇）
  • C 开源库之cJSON
  • 比特币8.12学习问题
  • 学懂C++（二十三）：高级教程——深入详解C++ 标准库的多线程支持
  • JavaScript 详解——Vue基础
  • vue 后台管理 之 状态管理 vuex 的使用
  • React之简易笔记本
  • k8s部署kubeadm init初始化不成功，coredns处于pending,master和nodes处于notready状态
  • MySQL 查询分组内最新的第一条数据
  • 〔开发系列〕一次关于小程序开发的深度总结
  • CentOS学习笔记 - 12. Nginx搭建Centos7.5远程repo
  • Elasticsearch 参考指南（升级前重新索引）
  • express + mock 让前后台并行开发
  • HashMap剖析之内部结构
  • javascript 总结（常用工具类的封装）
  • PHP面试之三：MySQL数据库
  • Python学习之路13-记分
  • quasar-framework cnodejs社区
  • React-flux杂记
  • v-if和v-for连用出现的问题
  • vue+element后台管理系统，从后端获取路由表，并正常渲染
  • 基于webpack 的 vue 多页架构
  • 聊聊spring cloud的LoadBalancerAutoConfiguration
  • 前端路由实现-history
  • 前端面试总结（at, md）
  • 一个JAVA程序员成长之路分享
  • 用Visual Studio开发以太坊智能合约
  • 最简单的无缝轮播
  • 【运维趟坑回忆录】vpc迁移 - 吃螃蟹之路
  • ​Python 3 新特性：类型注解
  • ​决定德拉瓦州地区版图的关键历史事件
  • ​学习笔记——动态路由——IS-IS中间系统到中间系统（报文/TLV）​
  • # 数论-逆元
  • #我与虚拟机的故事#连载20：周志明虚拟机第 3 版：到底值不值得买？
  • $().each和$.each的区别
  • （C#）if (this == null)？你在逗我，this 怎么可能为 null！用 IL 编译和反编译看穿一切
  • (day6) 319. 灯泡开关
  • （Matalb时序预测）PSO-BP粒子群算法优化BP神经网络的多维时序回归预测
  • (poj1.2.1)1970(筛选法模拟)
  • （二）换源+apt-get基础配置+搜狗拼音
  • （附源码）spring boot基于Java的电影院售票与管理系统毕业设计 011449
  • (四)docker：为mysql和java jar运行环境创建同一网络，容器互联
  • (转)自己动手搭建Nginx+memcache+xdebug+php运行环境绿色版 For windows版
  • *Algs4-1.5.25随机网格的倍率测试-(未读懂题)
  • .bat批处理（十）：从路径字符串中截取盘符、文件名、后缀名等信息
  • .net 7和core版 SignalR
  • .Net 基于.Net8开发的一个Asp.Net Core Webapi小型易用框架
  • .Net高阶异常处理第二篇~~ dump进阶之MiniDumpWriter
  • .NET教程 - 字符串 编码 正则表达式（String Encoding Regular Express）