Ubuntu 16.04开发CUDA程序入门(二)
来源:互联网 发布:bigbang大爆炸软件 编辑:程序博客网 时间:2024/05/18 20:47
CUDA编程模型相关笔记[1]
- 学习书籍:[1] 刘金硕等.基于CUDA的并行程序设计.科学出版社.2014
- CUDA编程模型:
- 如上图所示,一个完整的CUDA程序由主机代码和设备代码两部分组成。主机代码部分在主机端CPU上串行执行,是普通的C代码;设备代码部分在设备端GPU上并行执行,称为内核(kernel)。kernel函数不是一个完整的程序,而是任务中能被分解为并行执行的步骤的集合。CPU执行的串行程序负责kernel启动之前进行数据准备和设备初始化的工作,以及在kernel之间进行一些串行计算。GPU执行的并行部分是在被称为grid和block的两个层次并行中完成的,即每个kernel函数存在两个层次的并行:网格(grid)中的线程块(block)间并行和线程块中的线程(thread)间并行。
内核函数
一个完整的CUDA程序包括了在CPU端执行的串行代码和在GPU端执行的并行代码。CUDA程序的编写流程如下:
CUDA程序主要包括以下几个步骤:
初始化GPU
- CUDA程序首次调用Runtime函数时会初始化设备。初始化时,Runtime函数为系统中的每个设备建立一个上下文,该上下文作为设备的主要上下文,被应用的主机线程共享。
主机端本地数据的准备工作
- 所要准备的数据主要指待放入GPU上执行的计算任务,即函数的输入参数集。
为输入参数和输出参数分配显存空间
在显存中可以分配的空间有两种:线性存储器和CUDA数组。- 线性存储器指的是全局存储器,可以分配一维、二维或三维的线性存储空间。其中,一维数组的分配函数为
cudaMalloc()
,二维或三维的线性存储器分配函数为cudaMallocPitch()
和cudaMalloc3D()
。 - CUDA数组可以通过
cudaMalloc3DArray()
分配一维、二维或三维的CUDA数组,而cudaMallocArray()
一般用于分配二维CUDA数组。
- 线性存储器指的是全局存储器,可以分配一维、二维或三维的线性存储空间。其中,一维数组的分配函数为
将输入参数从主机端复制到显存
- 一维、二维、三维陷性存储器分别使用
cudaMemcpy()
、cudaMemcpy2D()
、cudaMemcpy3D()
进行主机端和设备端的数据传输。 - 由
cudaMalloc3DArray()
分配的CUDA数组使用cudaMemcpy3D()
完成与其他CUDA数组或者线性内存的数据传输。
- 一维、二维、三维陷性存储器分别使用
内核启动设置
- CUDA程序使用类似于
kernelFunc<<<blocksPerGrid, threadsPerBlock>>>(d_A, d_B, d_C);
的语句来启动kernelFunc
函数。其中,kernelFunc
为函数名,<<<>>>
运算符中的blocksPerGrid
和threadsPerBlock
是主机端告诉设备运行时如何启动kernelFunc
函数。blocksPerGrid
表示一个grid中有多少个并行block,threadsPerBlock
表示一个block中有多少个并行thread。(d_A, d_B, d_C)
为kernelFunc
函数的函数参数,和普通C函数一样。 - 在上一篇博客的代码中,执行内核的每个线程都会被分配一个独特的线程ID,可通过
blockDim.x * blockIdx.x + threadIdx.x
在内核中访问此ID(blockDim
指块的维度,类型为dim3;blockIdx
指网格内的块索引,类型为uint3;threadIdx
指块内的线程索引,类型为uint3)。在该程序中,GPU将自动产生blocksPerGrid * threadsPerBlock
个(即1*256个)ID不同的CUDA线程,相当于有256个执行加法命令的线程同时进行操作,从而总体时间缩短了256倍。
- CUDA程序使用类似于
将输入函数从显存复制到主机端
- 该步骤和将输入参数从主机端复制到显存几乎一样,唯一的区别在于,复制时的方向参数不同。将输入参数从主机端复制到显存的复制方向参数为
cudaMemcpyHostToDevice
,而将输入函数从显存复制到主机端的复制方向参数为cudaMemcpyDeviceToHost
。
- 该步骤和将输入参数从主机端复制到显存几乎一样,唯一的区别在于,复制时的方向参数不同。将输入参数从主机端复制到显存的复制方向参数为
释放在设备端分配的显存空间
- 释放线性存储器使用
cudaFree()
函数 - 释放CUDA数组使用
cudaFreeArray()
函数。
- 释放线性存储器使用
线程层次
- 在内核函数中,通过线程的索引来访问线程ID。在一维数组中,可以用一维线程块中的
threadIdx
直接指向相应的ID的线程,但是在二维数据、三维数据中却不相同。对于大小为(Dx,Dy)
的二维块,索引为(x,y)
的线程的ID是(x+yDx)
;对于大小为(Dx,Dy,Dz)
的三维块,索引为(x,y,z)
的线程的ID是(x+yDx+zDxDy)
。 - 针对不同维度的数组的索引方式,可以定义出二维、三维的线程块,去应对不同情况的数据并行方式。
0 0
- Ubuntu 16.04开发CUDA程序入门(二)
- Ubuntu 16.04开发CUDA程序入门(一)
- CUDA入门(二)cuda编程的基本知识与第一个cuda程序
- CUDA从入门到精通(二):第一个CUDA程序
- CUDA从入门到精通(二):第一个CUDA程序
- CUDA从入门到精通(二):第一个CUDA程序
- CUDA从入门到精通(二):第一个CUDA程序
- CUDA入门学习(二)
- Ubuntu下的CUDA编程(二)
- cuda入门之二
- CUDA程序优化小记(二)
- CUDA程序开发
- cuda入门:runtime API创建CUDA程序
- CUDA入门:runtime API创建CUDA程序
- 【CUDA开发】CUDA从入门到精通
- 《Java Web程序开发入门》知识总结(二)
- CUDA第一个程序优化二(有线程块)
- .NET Remoting程序开发入门篇(二)
- LintCode 二叉树的层次遍历
- C++之友元
- usb2.0及其它差分线布线时注意的问题
- add a response header on nginx when using proxy_pass
- audio和video元素
- Ubuntu 16.04开发CUDA程序入门(二)
- 高性能MySQL笔记-第一章
- 安装完固态硬盘后出现PCIIDE BAD的解决方法
- Python
- HDU 1856 More is better
- 2017年4月8日Office 365 使用CSV文件导入邮件组
- 记录windows驱动开发inf文件详解
- 数组替换——(问题还未解决)
- 替换元素