当前位置: 首页 > >

《GPU高性能编程CUDA实战》学*笔记 (1)

发布时间:

《GPU高性能编程CUDA实战》中代码整理
https://blog.csdn.net/fengbingchun/article/details/45954355
《GPU高性能编程CUDA实战》学*笔记(三)
https://blog.csdn.net/w09103419/article/details/52484969
CUDA编程??GPU架构,由sp,sm,thread,block,grid,warp说起
https://blog.csdn.net/junparadox/article/details/50540602
《CUDA并行程序设计-GPU编程指南》读书笔记?(1)线程网格、线程块以及线程
https://blog.csdn.net/u011475134/article/details/71577770


SP(streaming Process),SM(streaming multiprocessor)是硬件(GPU hardware)概念。而thread,block,grid,warp是软件上的(CUDA)概念。


SP:最基本的处理单元,streaming processor,也称为CUDA core。最后具体的指令和任务都是在SP上处理的。GPU进行并行计算,也就是很多个SP同时做处理。SM:多个SP加上其他的一些资源组成一个streaming multiprocessor。也叫GPU大核,其他资源如:warp scheduler,register,shared memory等。SM可以看做GPU的心脏(对比CPU核心),register和shared memory是SM的稀缺资源。CUDA将这些资源分配给所有驻留在SM中的threads。因此,这些有限的资源就使每个SM中active warps有非常严格的限制,也就限制了并行能力。

CUDA架构专门为GPU计算设计了一种全新的模块,目的是减轻早期GPU计算中存在的一些限制,而正是这些限制使得之前的GPU在通用计算中没有得到广泛的应用。
使用CUDA C来编写代码的前提条件包括
(1)、支持CUDA的图形处理器,即由NVIDIA推出的GPU显卡,要求显存超过256MB;
(2)、NVIDIA设备驱动程序,用于实现应用程序与支持CUDA的硬件之间的通信,确保安装最新的驱动程序,注意选择与开发环境相符的图形卡和操作系统;
(3)、CUDA开发工具箱即CUDA Toolkit,此工具箱中包括一个编译GPU代码的编译器;
(4)、标准C编译器,即CPU编译器。CUDA C应用程序将在两个不同的处理器上执行计算,因此需要两个编译器。其中一个编译器为GPU编译代码,而另一个为CPU编译代码。


第三章 第一段CUDA C代码+Host/Device
3.1 第一个程序
3.1.1 Hello World!

一般,将CPU以及系统的内存称为主机(Host),而将GPU及其内存称为设备(Device)。在GPU设备上执行的函数通常称为核函数(Kernel)。


#include
#include

__global__ void kernel(void) {
}

int main() {
kernel <<<1, 1>>>();
printf("hello world!");
getchar();
return 0;
}

这个程序与最初的“ Hello, World!”相比,多了两个值得注意的地方:


一个空的函数kernel(),并且带有修饰符__global__ 。这个修饰符将告诉编译器,函数应该编译为在设备而不是主机上运行。在这个简单的示例中,函数kernel()将被交给编译设备代码的编译器,而main()函数将被交给主机编译器。CUDA编译器和运行时将负责实现从主机代码中调用设备代码。

对这个空函数的调用,并且带有修饰字符 <<<1,1>>>


3.1.2 传递参数给核函数

#include
#include
#include

__global__ void add(int a, int b, int *c) {
*c = a + b;
}

int main()
{
int a = 2, b = 3, c = 0;
int* dev_c;
cudaMalloc((void**)&dev_c, sizeof(int));

//尖括号表示要将一些参数传递给CUDA编译器和运行时系统
//尖括号中这些参数并不是传递给设备代码的参数,而是告诉运行时如何启动设备代码,
//传递给设备代码本身的参数是放在圆括号中传递的,就像标准的函数调用一样
add << <1, 1 >> > (a, b, dev_c);
cudaMemcpy(&c, dev_c, sizeof(int), cudaMemcpyDeviceToHost);

printf("%d + %d = %d
", a, b, c);
cudaFree(dev_c);
getchar();
return 0;
}

注意这里增加了多行代码,在这些代码中包含两个概念:
? 可以像调用C函数那样将参数传递给核函数。
? 当设备执行任何有用的操作时,都需要分配内存,例如将计算值返回给主机。


通过cudaMalloc() 来分配内存。
这个函数调用的行为非常类似于标准的C函数malloc(),但该函数的作用是告诉CUDA运行时在设备上分配内存。


第一个参数是一个指针,指向用于保存新分配内存地址的变量,第二个参数是分配内存的大小。除了分配内存的指针不是作为函数的返回值外,这个函数的行为与malloc()是相同的,并且返回类型为void*。

cudaMalloc函数使用限制总结:
1.可以将cudaMalloc()分配的指针传递给在设备上执行的函数。
2.可以在设备代码中使用cudaMalloc()分配的指针进行内存读/写操作。
3.可以将cudaMalloc()分配的指针传递给在主机上执行的函数。
4.不能在主机代码中使用cudaMalloc()分配的指针进行内存读/写操作。
要释放cudaMalloc()分配的内存,需要调用cudaFree()
设备指针的使用方式与标准C中指针的使用方式完全一样。 主机指针只能访问主机代码中的内存,而设备指针也只能访问设备代码中的内存。


3.1.3 查询设备

有可能在单块卡*肆礁龌蚨喔鯣PU。
在主机代码中可以通过调用cudaMemcpy() 来访问设备上的内存,在设备和主机之间复制数据。
获得CUDA设备的数量,可以调用cudaGetDeviceCount()


int count;
HANDLE_ERROR( cudaGetDeviceCount( &count ) );

在集成的GPU上运行代码,可以与CPU共享内存。
计算功能集的版本为1.3或者更高的显卡才能支持双精度浮点数的计算。


3.1.4 设备属性的使用

cudaGetDeviceProperties()
查找主版本号大于1,或者主版本号为1且次版本号大于等于3的设备:
首先,找出我们希望设备拥有的属性并将这些属性填充到一个cudaDeviceProp结构。


cudaDeviceProp prop;
memset( &prop, 0, sizeof( cudaDeviceProp ) );
prop.major = 1;
prop.minor = 3;

在填充完 cudaDeviceProp 结构后,将其传递给 cudaChooseDevice(),这样CUDA运行时将查找是否存在某个设备满足这些条件。 cudaChooseDevice()函数将返回一个设备ID,然后我们可以将这个ID传递给 cudaSetDevice()。随后,所有的设备操作都将在这个设备上执行。


#include "../common/book.h"
int main( void )
{
cudaDeviceProp prop;
int dev;
HANDLE_ERROR( cudaGetDevice( &dev ) );
printf( "ID of current CUDA device: %d
", dev );
memset( &prop, 0, sizeof( cudaDeviceProp ) );
prop.major = 1;
prop.minor = 3;
HANDLE_ERROR( cudaChooseDevice( &dev, &prop ) );
printf( "ID of CUDA device closest to revision 1.3: %d
", dev );
HANDLE_ERROR( cudaSetDevice( dev ) );
}



友情链接: