CUDA并行程序设计:Hello world!

[!] 因个人水平问题,以下内容可能会出现部分错误的情况,还望各位批评指正



文章目录

  • CUDA并行程序设计:Hello world!
  • 1. 什么是CUDA?
  • 2. GPU和CPU的区别在哪里?
  • 3. GPU中的线程管理
  • 4. 用VS2017写一个Hello world!
  • 4.1 创建新工程
  • 4.2 具体代码
  • 4.3 运行结果
  • 4.4 性能比较
  • 5 总体代码(太长不看版)



1. 什么是CUDA?

  • 2007年,英伟达NVIDIA为了让GPU进入主流,为GPU增加了一个易用的编程接口,称为统一计算架构 (Compute Unified Device Architecture, CUDA) 。CUDA是C语言的一种扩展,它能够使用标注C来进行GPU的代码编程,所以CUDA既能够用于在CPU上运行,也适用于在GPU上运行。

2. GPU和CPU的区别在哪里?

  • CPU作为计算机系统的运算和控制核心,是信息处理、程序运行的最终执行单元。因此他的通用性比较强,能够运行许多足够复杂的计算,但是计算能力比较慢。
  • GPU则存在着大量的计算核心,计算能力非常优秀,但是只能运行简单粗暴的计算。

如何计算GPU的并发量 gpu并行计算与cuda编程_c语言


[网图侵删]

3. GPU中的线程管理

  • CUDA是基于SPMD模型(Single Program Multiple Data, 单程序多数据)来处理代码的,CUDA要求程序员都要把问题分成若干个小块来处理,每一个块送到GPU中的一个处理单元上处理。在一个具体的代码中,GPU的被分为线程块线程束两类,32个线程组成一个线程束,若干个线程束组成一个线程块。为了提高效率,线程块中的线程数通常是32的倍数,也就是有整数倍个线程束。
  • 在CUDA中,我们可以通过创建一个内核函数来实现并行化,所谓的内核函数,就是一个只能在GPU上执行而不能直接在CPU上执行的函数,具体长这个样子。
__global__ void function(parm1,parm2,...)
{
	//代码块
}

4. 用VS2017写一个Hello world!

本处默认大家都成功安装好CUDA和Visual Studio。至于安装过程已经有很多帖子有描述了,所以这里就不再复述。本文CUDA版本为10.1,Visual Studio为2017,但理论上所有CUDA版本和VS版本都能够运行。

4.1 创建新工程

  1. 首先,打开“新建项目”,找到"NVIDIA"项,创建一个"CUDA X.Y Runtime"项目(其中X.Y是CUDA版本号)。
  2. 如何计算GPU的并发量 gpu并行计算与cuda编程_c语言_02

  3. 创建后,会发现里面内置了一个 kernel.cu 代码,这个代码也是一个完整的示例代码,可以直接运行。如果运行成功了,就代表CUDA安装并没有问题。
  4. 如何计算GPU的并发量 gpu并行计算与cuda编程_CUDA_03

  5. 直接把kernel.cu所有内容清空掉,来准备我们的Hello world程序。

4.2 具体代码

  1. 导入项目所需要的库
//CUDA所需要的库
#include "cuda_runtime.h"
#include "device_launch_parameters.h"

//C语言标准库
#include <string.h>
#include <stdio.h>
#include <time.h>
  1. 构建一个内核函数,该函数的功能为:显示自己的线程块和线程ID,并输出一句话
//内核函数,用于显示自己的线程块ID和线程ID,并输出string字符串中的内容
__global__ void cudaPrintf(char * string)
{
	printf("Block %d - Thread %d: %s\n", blockIdx.x, threadIdx.x, string);
}
  1. 构建主函数,并初始化一些参数
int main(void)
{
	clock_t start = clock();  //记录运行时间
	register int blockNum = 4; //线程块数量
	register int threadNum = 32;  //每个块内的线程数量
	char * string = "Hello world!";  //需要显示的内容,本文为Hello world!
	char * gpuString = NULL;       //需要输入到GPU中的字符串
}
  1. 在GPU中申请空间来存放字符串,并且将显示内容从CPU搬运到GPU中
//在GPU中申请足够的空间存放字符串
cudaMalloc((void **)&gpuString, sizeof(char) * strlen(string)); 

//将CPU中的 string字符串复制到GPU中的gpuString变量上
cudaMemcpy((char * )gpuString, string, sizeof(char) * strlen(string),cudaMemcpyHostToDevice);
  1. 运行核函数
//其中cudaPrintf为函数名 <<<线程块,线程数>>> 表示了需要调用多少个线程块和线程束来运行
cudaPrintf<<<blockNum, threadNum>>>(gpuString);
  1. 计算总时间
printf("total time: %fs\n", (double)(clock() - start) / CLOCKS_PER_SEC);
  1. 释放CUDA内存
cudaFree(gpuString);

4.3 运行结果

  • 代码运行结果如下

4.4 性能比较

  • 在笔者电脑中,当选择的代码块数量为4096,每个块的线程数为32时,运算的时间为:4.260秒
  • 在笔者电脑中,当直接循环4096 * 32次时,所需要时间为:14.154秒
  • 两者相比,使用GPU所需要的时间为CPU的0.300975倍,即大约快了三倍以上。
  • 笔者电脑CPU为 i7-9750H,GPU为RTX 2070。

5 总体代码(太长不看版)

  • 总体代码如下
#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <string.h>
#include <stdio.h>
#include <time.h>

__global__ void cudaPrintf(char * string)
{
  printf("Block %d - Thread %d: %s\n", blockIdx.x, threadIdx.x,string);
}

int main(void)
{
  clock_t start = clock();
  register int blockNum = 1024;
  register int threadNum = 32;
  char * string = "Hello world!";
  char * gpuString = NULL;

  cudaMalloc((void **)&gpuString, sizeof(char) * strlen(string));
  cudaMemcpy((char * )gpuString, string, sizeof(char) * strlen(string),cudaMemcpyHostToDevice);
  cudaPrintf<<<blockNum, threadNum>>>(gpuString);
  
  cudaFree(gpuString);
  printf("total time: %fs\n", (double)(clock() - start) / CLOCKS_PER_SEC);
  return 0;
}