CUDA入门

article/2025/9/4 22:28:51

1. 引言

CUDA为a platform and programming model for CUDA-enabled GPUs。该平台通过GPU来进行计算。CUDA为GPU编程和管理 提供C/C++语言扩展和API。

CUDA编程中，会同时使用CPU和GPU进行计算：

  • CPU system：称为host。
  • GPU system：称为device。

CPU和GPU为独立的系统，具有各自的内存空间。通常，在CPU上运行的串行工作，而将并行计算卸载给GPU。

2. CUDA和C对比

以Hello world程序为例：
在这里插入图片描述
二者最大的不同在于__global__说明符 和 <<<...>>>语法：

  • __global__说明符：用于标明该函数运行于device（GPU）。这类函数可通过host code调用，如通过main()函数调用。也可被称为“kernels”。
  • <<<...>>>语法：当kernel被调用时，其执行配置由<<<...>>>语言提供，如cuda_helo<<<1,1>>>()。在CUDA术语中，这被称为“kernel launch”。

编译CUDA程序与编译C语言类似。NVIDIA在其CUDA toolkit中提供了名为nvcc的CUDA编译器来编译CUDA code——通常源代码文件名为.cu。

以vector addition为例，相应的C语言实现为（vector_add.c）：

#define N 10000000void vector_add(float *out, float *a, float *b, int n) {for(int i = 0; i < n; i++){out[i] = a[i] + b[i];}
}int main(){float *a, *b, *out; // Allocate memorya   = (float*)malloc(sizeof(float) * N);b   = (float*)malloc(sizeof(float) * N);out = (float*)malloc(sizeof(float) * N);// Initialize arrayfor(int i = 0; i < N; i++){a[i] = 1.0f; b[i] = 2.0f;}// Main functionvector_add(out, a, b, N);
}

对应的CUDA程序（vector_add.cu）若为：

#define N 10000000__global__  void vector_add(float *out, float *a, float *b, int n) {for(int i = 0; i < n; i++){out[i] = a[i] + b[i];}
}int main(){float *a, *b, *out; // Allocate memorya   = (float*)malloc(sizeof(float) * N);b   = (float*)malloc(sizeof(float) * N);out = (float*)malloc(sizeof(float) * N);// Initialize arrayfor(int i = 0; i < N; i++){a[i] = 1.0f; b[i] = 2.0f;}// Main functionvector_add<<<1,1>>>(out, a, b, N);
}

但是，以上CUDA程序并无法运行，因为CPU和GPU为不同的实体，二者具有各自的内存空间。CPU无法直接访问GPU内存，GPU也无法直接访问CPU内存。在CUDA术语中：

  • CPU内存：称为host memory。指向CPU内存的指针称为host pointer。
  • GPU内存：称为device memory。指向GPU内存的指针称为device pointer。

GPU想访问的数据因存储在device memory中。CUDA提供了API来分配device memory，并进行host memory和device memory之间的数据传送。

CUDA程序的基本流程为：

  • 1）分配host memory并初始化host data。
  • 2）分配device memory：可通过cudaMalloc()和cudaFree()来分配和释放device memory。分别于C语言中的malloc()和free()对应。
cudaMalloc(void **devPtr, size_t count);
cudaFree(void *devPtr);
  • 3）内存转换：将host data由host memory 传送至 device memory。可通过cudaMemcpy函数来实现数据拷贝，类似于C语言中的memcpy。
// kind表示方向，为`cudaMemcpyHostToDevice` 或 `cudaMemcpyDeviceToHost`
cudaMemcpy(void *dst, void *src, size_t count, cudaMemcpyKind kind)
  • 4）运行kernels。
  • 5）将device memory执行的结果输出到host memory。

根据以上流程，对应的CUDA程序（vector_add.cu）需修改为：

#include <stdio.h>
#include <stdlib.h>
#include <math.h>
#include <assert.h>
#include <cuda.h>
#include <cuda_runtime.h>#define N 10000000
#define MAX_ERR 1e-6__global__ void vector_add(float *out, float *a, float *b, int n) {for(int i = 0; i < n; i ++){out[i] = a[i] + b[i];}
}int main(){float *a, *b, *out;float *d_a, *d_b, *d_out; // Allocate host memorya   = (float*)malloc(sizeof(float) * N);b   = (float*)malloc(sizeof(float) * N);out = (float*)malloc(sizeof(float) * N);// Initialize host arraysfor(int i = 0; i < N; i++){a[i] = 1.0f;b[i] = 2.0f;}// Allocate device memorycudaMalloc((void**)&d_a, sizeof(float) * N);cudaMalloc((void**)&d_b, sizeof(float) * N);cudaMalloc((void**)&d_out, sizeof(float) * N);// Transfer data from host to device memorycudaMemcpy(d_a, a, sizeof(float) * N, cudaMemcpyHostToDevice);cudaMemcpy(d_b, b, sizeof(float) * N, cudaMemcpyHostToDevice);// Executing kernel vector_add<<<1,1>>>(d_out, d_a, d_b, N);// Transfer data back to host memorycudaMemcpy(out, d_out, sizeof(float) * N, cudaMemcpyDeviceToHost);// Verificationfor(int i = 0; i < N; i++){assert(fabs(out[i] - a[i] - b[i]) < MAX_ERR);}printf("PASSED\n");// Deallocate device memorycudaFree(d_a);cudaFree(d_b);cudaFree(d_out);// Deallocate host memoryfree(a); free(b); free(out);
}

编译并可使用time来验证程序性能：

$> nvcc vector_add.cu -o vector_add
$> time ./vector_add

NVIDIA也提供了名为nvprof的命令行profiler工具，可提供更多程序性能信息：

$> nvprof ./vector_add

以Tesla M2050为例，相应profiling为：

==6326== Profiling application: ./vector_add
==6326== Profiling result:
Time(%)      Time     Calls       Avg       Min       Max  Name97.55%  1.42529s         1  1.42529s  1.42529s  1.42529s  vector_add(float*, float*, float*, int)1.39%  20.318ms         2  10.159ms  10.126ms  10.192ms  [CUDA memcpy HtoD]1.06%  15.549ms         1  15.549ms  15.549ms  15.549ms  [CUDA memcpy DtoH]

3. CUDA并行化

CUDA使用kernel execution configuration <<<...>>> 来告诉CUDA runtime该在GPU中启动多少个线程。
CUDA organizes threads into a group called “thread block”。
Kernel可启动多个thread blocks，organized into a “grid” structure。

kernel execution configuration的语法为：

<<<M, T>>>

表示kernel launches with a gird of M thread blocks。每个thread block具有T parallel threads。

接下来，将使用multithread来parallelize上例中的vector addition，如使用a thread block with 256 threads，相应的kernel execution configuration为：

vector_add <<< 1 , 256 >>> (d_out, d_a, d_b, N);

CUDA提供了内置变量来访问thread information，此例中包含了一下2个内置变量：

  • threadIdx.x：包含了the index of the thread within the block。此例中，index范围为0~255。
  • blockDim.x：包含了the size of thread block（number of threads in the thread block）。此例中，该值为256。

在这里插入图片描述

完整的vector_add_thread.cu源代码为：

#include <stdio.h>
#include <stdlib.h>
#include <math.h>
#include <assert.h>
#include <cuda.h>
#include <cuda_runtime.h>#define N 10000000
#define MAX_ERR 1e-6__global__ void vector_add(float *out, float *a, float *b, int n) {int index = threadIdx.x;int stride = blockDim.x;for(int i = index; i < n; i += stride){out[i] = a[i] + b[i];}
}int main(){float *a, *b, *out;float *d_a, *d_b, *d_out; // Allocate host memorya   = (float*)malloc(sizeof(float) * N);b   = (float*)malloc(sizeof(float) * N);out = (float*)malloc(sizeof(float) * N);// Initialize host arraysfor(int i = 0; i < N; i++){a[i] = 1.0f;b[i] = 2.0f;}// Allocate device memory cudaMalloc((void**)&d_a, sizeof(float) * N);cudaMalloc((void**)&d_b, sizeof(float) * N);cudaMalloc((void**)&d_out, sizeof(float) * N);// Transfer data from host to device memorycudaMemcpy(d_a, a, sizeof(float) * N, cudaMemcpyHostToDevice);cudaMemcpy(d_b, b, sizeof(float) * N, cudaMemcpyHostToDevice);// Executing kernel vector_add<<<1,256>>>(d_out, d_a, d_b, N);// Transfer data back to host memorycudaMemcpy(out, d_out, sizeof(float) * N, cudaMemcpyDeviceToHost);// Verificationfor(int i = 0; i < N; i++){assert(fabs(out[i] - a[i] - b[i]) < MAX_ERR);}printf("PASSED\n");// Deallocate device memorycudaFree(d_a);cudaFree(d_b);cudaFree(d_out);// Deallocate host memoryfree(a); free(b); free(out);
}
$> nvcc vector_add_thread.cu -o vector_add_thread
$> nvprof ./vector_add_thread

相应的性能为：

==6430== Profiling application: ./vector_add_thread
==6430== Profiling result:
Time(%)      Time     Calls       Avg       Min       Max  Name39.18%  22.780ms         1  22.780ms  22.780ms  22.780ms  vector_add(float*, float*, float*, int)34.93%  20.310ms         2  10.155ms  10.137ms  10.173ms  [CUDA memcpy HtoD]25.89%  15.055ms         1  15.055ms  15.055ms  15.055ms  [CUDA memcpy DtoH]

以上为1个thread block。CUDA GPU具有多个并行处理器，名为Streaming Multiprocessors（SMs）。每个SM包含了多个并行处理器，可运行多个concurrent thread blocks。为了充分利用CUDA GPU，kernel应启动多个thread blocks。此时CUDA再额外提供2个内置变量：

  • blockIdx.x：包含the index of the block with in the grid。
  • gridDim.x：包含the size of the grid。

在这里插入图片描述
若一共需要 N N N个线程，每个thread block有256个线程，则至少需要 N / 256 N/256 N/256个thread blocks。对于每个thread需要有a unique index，该index的计算规则为：

int tid = blockIdx.x * blockDim.x + threadIdx.x;

多个thread block的vector_add_grid.cu源代码为：

#include <stdio.h>
#include <stdlib.h>
#include <math.h>
#include <assert.h>
#include <cuda.h>
#include <cuda_runtime.h>#define N 10000000
#define MAX_ERR 1e-6__global__ void vector_add(float *out, float *a, float *b, int n) {int tid = blockIdx.x * blockDim.x + threadIdx.x;// Handling arbitrary vector sizeif (tid < n){out[tid] = a[tid] + b[tid];}
}int main(){float *a, *b, *out;float *d_a, *d_b, *d_out; // Allocate host memorya   = (float*)malloc(sizeof(float) * N);b   = (float*)malloc(sizeof(float) * N);out = (float*)malloc(sizeof(float) * N);// Initialize host arraysfor(int i = 0; i < N; i++){a[i] = 1.0f;b[i] = 2.0f;}// Allocate device memory cudaMalloc((void**)&d_a, sizeof(float) * N);cudaMalloc((void**)&d_b, sizeof(float) * N);cudaMalloc((void**)&d_out, sizeof(float) * N);// Transfer data from host to device memorycudaMemcpy(d_a, a, sizeof(float) * N, cudaMemcpyHostToDevice);cudaMemcpy(d_b, b, sizeof(float) * N, cudaMemcpyHostToDevice);// Executing kernel int block_size = 256;int grid_size = ((N + block_size) / block_size);vector_add<<<grid_size,block_size>>>(d_out, d_a, d_b, N);// Transfer data back to host memorycudaMemcpy(out, d_out, sizeof(float) * N, cudaMemcpyDeviceToHost);// Verificationfor(int i = 0; i < N; i++){assert(fabs(out[i] - a[i] - b[i]) < MAX_ERR);}printf("PASSED\n");// Deallocate device memorycudaFree(d_a);cudaFree(d_b);cudaFree(d_out);// Deallocate host memoryfree(a); free(b); free(out);
}

编译并profile性能：

$> nvcc vector_add_grid.cu -o vector_add_grid
$> nvprof ./vector_add_grid

在Tesla M2050上的性能表现为：

==6564== Profiling application: ./vector_add_grid
==6564== Profiling result:
Time(%)      Time     Calls       Avg       Min       Max  Name55.65%  20.312ms         2  10.156ms  10.150ms  10.162ms  [CUDA memcpy HtoD]41.24%  15.050ms         1  15.050ms  15.050ms  15.050ms  [CUDA memcpy DtoH]3.11%  1.1347ms         1  1.1347ms  1.1347ms  1.1347ms  vector_add(float*, float*, float*, int)

4. 性能对比

VersionExecution Time (ms)Speedup
1 thread1425.291.00x
1 block22.7862.56x
Multiple blocks1.131261.32x

5. OpenCL

OpenCL全称为：Open Computing Language。
OpenCL为：

  • Open, royalty-free standard C-language extension
  • For parallel programming of heterogeneous system using GPUs, CPUS, CBE, DSP’s and other processors including embedded mobile devices。
  • 初始由苹果公司发起。苹果公司put OpenCL in OSX Snow Leopard and is active in the working group。Wroking group内包含NVIDIA， Intel，AMD，IBM等等。
  • 由Khronos Group管理。该Group同时管理了OpenGL std。

在这里插入图片描述
基本的程序结构为：
在这里插入图片描述

参考资料

[1] Getting started with OpenCL and GPU Computing
[2] Introduction to GPU Computing with OpenCL
[3] OpenCL™ Programming Guide for the CUDA™ Architecture
[4] CUDA Tutorial


http://chatgpt.dhexx.cn/article/xkIKpHtF.shtml

相关文章

cuda和cudatoolkit

cuda和cudatoolkit

Pytorch 使用不同版本的 cuda 由于课题的原因&#xff0c;笔者主要通过 Pytorch 框架进行深度学习相关的学习和实验。在运行和学习网络上的 Pytorch 应用代码的过程中&#xff0c;不少项目会标注作者在运行和实验时所使用的 Pytorch 和 cuda 版本信息。由于 Pytorch 和 cuda 版…
阅读更多...
最新CUDA环境配置(Win10 + CUDA 11.6 + VS2019)

最新CUDA环境配置(Win10 + CUDA 11.6 + VS2019)

最新CUDA环境配置(Win10 CUDA 11.6 VS2019) 本篇博客根据NVIDIA 官方文档所述, 并根据自己实践得出. 供各位需要的朋友参考. 1.前言 本篇文章的软件环境为: Windows 10CUDA 11.6VS2019 CUDA是目前做人工智能, 深度学习等方向的必备工具库. 由CUDA衍生出的加速工具很多, …
阅读更多...
一文搞懂CUDA

一文搞懂CUDA

什么是cuda 统一计算设备架构&#xff08;Compute Unified Device Architecture, CUDA&#xff09;&#xff0c;是由NVIDIA推出的通用并行计算架构。解决的是用更加廉价的设备资源&#xff0c;实现更高效的并行计算。 CUDA是NVIDIA公司所开发的GPU编程模型&#xff0c;它提供…
阅读更多...
GPU，CUDA，cuDNN的理解

GPU，CUDA，cuDNN的理解

我们知道做深度学习离不开GPU，不过一直以来对GPU和CPU的差别，CUDA以及cuDNN都不是很了解，所以找了些资料整理下，希望不仅可以帮助自己理解，也能够帮助到其他人理解。 先来讲讲CPU和GPU的关系和差别吧。截图来自资料1（CUDA的官方文档）： 从上图可以看出GPU（图像处理器，…
阅读更多...
CUDA编程之快速入门

CUDA编程之快速入门

CUDA（Compute Unified Device Architecture）的中文全称为计算统一设备架构。做图像视觉领域的同学多多少少都会接触到CUDA，毕竟要做性能速度优化，CUDA是个很重要的工具，CUDA是做视觉的同学难以绕过的一个坑，必须踩一踩才踏实。CUDA编程真的是入门容易精通难，具有计算机体…
阅读更多...
CUDA学习

CUDA学习

想想学习CUDA的时间也应该有十来天了&#xff0c;也该是做一个小总结了&#xff0c;说说我理解的CUDA&#xff0c;它到底是什么东西&#xff1f; 其实说到CUDA&#xff0c;还真的没几个人知道&#xff0c;说实话&#xff0c;我也听说不久&#xff0c;主要因为它2007年才刚发布&…
阅读更多...
CUDA简介

CUDA简介

CUDA简介 CUDA是什么 CUDA&#xff0c;Compute Unified Device Architecture的简称&#xff0c;是由NVIDIA公司创立的基于他们公司生产的图形处理器GPUs&#xff08;Graphics Processing Units,可以通俗的理解为显卡&#xff09;的一个并行计算平台和编程模型。 通过CUDA&#…
阅读更多...
CUDA是什么-CUDA简介

CUDA是什么-CUDA简介

在大家开始深度学习时&#xff0c;几乎所有的入门教程都会提到CUDA这个词。那么什么是CUDA&#xff1f;她和我们进行深度学习的环境部署等有什么关系&#xff1f;通过查阅资料&#xff0c;我整理了这份简洁版CUDA入门文档&#xff0c;希望能帮助大家用最快的时间尽可能清晰的了…
阅读更多...
java队列和栈 共同_java 栈和队列的模拟--java

java队列和栈 共同_java 栈和队列的模拟--java

栈的定义&#xff1a;栈是一种特殊的表这种表只在表头进行插入和删除操作。因此&#xff0c;表头对于栈来说具有特殊的意义&#xff0c;称为栈顶。相应地&#xff0c;表尾称为栈底。不含任何元素的栈称为空栈。 栈的逻辑结构&#xff1a;假设一个栈S中的元素为an,an-1,..,a1&am…
阅读更多...
栈和队列学习总结

栈和队列学习总结

一、栈 1、特点及应用 先进后出。（如果会和队列先进先出记混的话，就记场景吧：弹栈弹栈，就是把最上面的最新进来的弹出去；而队列就像我们火车站排队检票出站一样，谁排在前面谁就先出去。） 应用的话，其实我们经常接触呀。比如Undo操作（就是撤销操作）就是使用的栈的思想…
阅读更多...
栈和队列的共同点和不同点

栈和队列的共同点和不同点

堆栈都是一种数据项按序排列的数据结构&#xff0c;只能在一端(称为栈顶(top))对数据项进行插入和删除。 要点&#xff1a;堆&#xff1a;顺序随意 栈&#xff1a;后进先出(Last-In/First-Out) 堆 堆&#xff1a;什么是堆&#xff1f;又该怎么理解呢&#xff1f; ①堆通常是一…
阅读更多...
栈和队列实现和实例分析

栈和队列实现和实例分析

目录 前言栈队列实例分析结语 前言 本篇文章主要讲述数据结构中栈和队列的实现&#xff0c;以及相关实例分析。 栈 注意本文所讲述的栈是数据结构的一种&#xff0c;并不是内存划区中的栈区&#xff0c;但是这两者有相似之处&#xff0c;即&#xff1a;存储数据时满足数据先…
阅读更多...
栈和队列的共同处和不同处

栈和队列的共同处和不同处

共同处 栈和队列的共同处是&#xff1a;它们都是由几个数据特性相同的元素组成的有限序列&#xff0c;也就是所谓的线性表。 不同处 队列 队列&#xff08;queue&#xff09;是限定仅在表的一端插入元素、在另一端删除元素的线性表。 在队列中&#xff0c;允许插入的一端被…
阅读更多...
索引的优缺点以及索引的设计原则

索引的优缺点以及索引的设计原则

索引概述 索引&#xff08;index&#xff09; 是帮助 MySQL 高效获取数据的数据结构&#xff08;有序&#xff09;。 在数据之外&#xff0c;数据库系统还维护者满足特定查找算法的数据结构&#xff0c;这些数据结构以某种方式引用&#xff08;指向&#xff09;数据&#xff0…
阅读更多...
SQL数据库之索引优缺点

SQL数据库之索引优缺点

SQL数据库之索引使用原则及利弊 索引是对数据库表中一列或多列的值进行排序的一种结构&#xff0c;使用索引可快速访问数据库表中的特定信息。 优点 通过创建唯一性索引&#xff0c;可以保证数据库表中每一行数据的唯一性。 可以大大加快数据的检索速度&#xff0c;这也是创建…
阅读更多...
Oracle索引的建立及优缺点

Oracle索引的建立及优缺点

在看公司建表语句时发现了这样一段代码 本着学习的态度面向百度&#xff1a;&#xff1a;&#xff1a;&#xff1a;&#xff1a; 原来这是Oracle的索引 Oracle的索引说明 1&#xff09;索引是数据库对象之一&#xff0c;用于加快数据的检索&#xff0c;类似于书籍的索引。在…
阅读更多...
MySQL索引的优缺点

MySQL索引的优缺点

MySQL 中的索引简介 1、索引的优点 为什么要创建索引&#xff1f;这是因为&#xff0c;创建索引可以大大提高系统的查询性能。 第一、通过创建唯一性索引&#xff0c;可以保证数据库表中每一行数据的唯一性。 第二、可以大大加快数据的检索速度&#xff0c;这也是创建索引的最…
阅读更多...
MySQL索引的使用知识有哪些？

MySQL索引的使用知识有哪些？

面试造火箭&#xff0c;工作拧螺丝&#xff0c;虽然工作时我们都在使用基本的 sql&#xff0c;但是不好意思&#xff0c;面试 90% 都在问原理&#xff0c;例如索引&#xff0c;锁&#xff0c;日志&#xff0c;引擎啊等等。 在关系数据库中&#xff0c;索引是一种单独的、物理的…
阅读更多...
索引的数据结构与优缺点

索引的数据结构与优缺点

1、索引的数据结构 什么是索引&#xff1f; 索引就是mysql为了提高查询数据的一种数据结构。在数据之外&#xff0c;数据库系统还维护着满足特定查找算法 的数据结构&#xff0c;这些数据结构以某种方式引用(指向)数据&#xff0c;这样就可以在这些数据结构上实现高级查找 算法…
阅读更多...
Oracle 数据库：ORA-12541: TNS: 无监听程序 的解决办法

Oracle 数据库：ORA-12541: TNS: 无监听程序 的解决办法

路径下D:\app\ou\product\11.2.0\dbhome_1\NETWORK\ADMIN host后面地址全改为 " HOSTlocalhost " 。 以上两个配置文件修改完成后&#xff0c;Win R 在弹出框中输入 " SERVICES.MSC " &#xff0c;找到Oracle的服务&#xff08;OracleService&#xff09;…
阅读更多...
推荐文章

Copyright @ 2022~2023