CUDA编程接口（二）------一十八般武器------GPU的革命

CUDA编程接口（二）------一十八般武器

------GPU的革命

<!--[if !supportLists]-->4. <!--[endif]-->程序运行控制：像Stream，Event，Context, Module, Execution control这样的咱都把归类到运行管理的。这里也得分清楚有Runtime级别的，也有Driver级别的。

Stream：如果了解AGP时代的显卡，就知道在Device和Host之间进行数据交换的时候有作为中转数据的部分，叫着stream；发展到G8X以后就有了新的专门针对GPGPU统一设计的方便数据传输的Stream out的硬件设备。这个层的功能是将Vertex Shader和Pixel Shader（在G8X的设计中已经差不多不用把他们俩再分开，差不多统一起来）处理完成的数据输出给用户，由用户进行处理后再反馈给流水线继续处理。它可以直接读写本地显存。如果是内存对齐的，速度会更快，这个部分可以参考simpleStream的代码。

Event：和平时的大家在编程中使用的Event一个道理，都是起到通知的作用，创建一个事件，然后同步；Event在多线程编程中会经常遇到。

Context：上下文，什么叫叫上下文- -！（中文翻译叫上下文，实在的，俺觉得这个翻译不恰当，但是很多传统的书都这么翻译了，咱们也只能叫上下文）这里的Context和CPU里面使用的Context差不多，都是一个process里面需要用到的一些“资源”（系统资源：堆栈，内存……等等）。我倒觉得翻译为process包含的东东更为恰当。

Module：可以理解为linux里面的module意思；如果你不了解linux也不用担心，module就是一个个的专门针对Device的程序。知道很久很久以前的dos时代就有的.com文件吗？那个东东就像直接在CPU上可以调用的一个文件，加载到内存中就可以运行的。这里的module就是加载到Device上就可以运行。

Execution control：这个就是从驱动层面来控制线程怎么在Device上运行。

Stream和Event在Runtime API和Driver API中都有，函数接口也差不多。

Context, Module, Execution control就只有在Driver API层面才有。

上面的这些都没多少难点，其实API的讲解，一般都没多少难点，难点在于怎么能灵活的运用API，那就得通过多次的练习，多用API实践；其实有的时候也能发现API的不足，才会产生新的API。这里就不放具体的代码，代码可以参考编程手册上面。

<!--[if !supportLists]-->5. <!--[endif]-->好了，这里就剩下OpenGL和Direct3D的接口函数了，他也有两个层次的API，有Runtime层次的，也有Driver层次的。Runtime和Driver层面都有调用的API，CUDA2.0在这里也做了一些优化，在内存交换方面。具体代码参考编程手册。

前面说过要讲解函数的定义，其实在以前的帖子里面正好已经翻译过一段，不过现在要讲解，也就提出来，专门再把那些东西翻出来：GPU NVIDIA_CUDA_Programming_Cuide_1.0 Chapter 4. Application Programming Interface

4.2.1 Function Type Qualifiers 函数类型

4.2.1.1 __device__

__device__ 规定的函数：

在device上执行。

只能在device上调用。

4.2.1.2 __global__

__global__ 定义了一个kernel函数：

在device上运行。

只能在host上调用

4.2.1.3 __host__

__host__定义的函数：

在host上运行，

只能在host上调用。

没有定义__host__，__device__或者__global__的函数等同于__host__函数，系统都会把函数编译成host函数。

另外，__host__定义可以和__device__定义一起使用，编译器会把这个函数编译为host和device通用的函数。

4.2.1.4 Restrictions (强调，限制)

__device__函数通常是inline，所以如果需要不inline 就需要加上uninline

__device__和__global__函数都不支持递归调用。

__device__和__global__函数都不能定义static变量在函数内部。

__device__和__global__函数不能使用可变参数。

__device__函数没有函数地址，也没有指向它的函数指针，但是__global__函数有。

__global__定义和__host__定义不能一起使用。

__global__函数必须是void返回类型。

任何调用__global__的函数都必须指明运行配置。（Section 4.2.3）

也就是线程kernel的调用方法。

__global__函数是异步调用的。在运行结束前就会返回。

__global__函数的参数通常是通过shared memory调用到Device中，最多是265bytes。