nVidia CUDA API（下）

 

 

在《nVidia CUDA API（上）》的部分，已经大致说明了 extension 的部分；这边要来讲的则是 runtime library 的部分。

CUDA 的 runtime library 分成下面三部分：

• common component
  提供 CUDA 的 vector 等型别，以及可以同时在 host 及 device 上可以执行的函式。主要包括：

  • 内建的 vector 型别

  • dim3 型别（用於指定 kernel 参数的型别）

  • texture 型别

  • 数学函式

• device component
  在 device 上执行的部分，提供 device 上特殊的函式。主要有：

  • 数学函式。这边的数学函式是 device 的特殊版本，精确度较低，但是速度较快。

  • 同步函式

  • Atomic 函式

  • 型别转换函式（Conversion/Casting）

  • texture 函式

• host component
  在 host 上执行的部分，负责控制 device。主要是处理：

  • Device 管理

  • Context 管理（所谓 Context 大致相当於 CPU？）

  • 记忆体管理

  • Code module 管理（Code module 似乎相当於 dynamic library）

  • Execution control

  • Texture reference 管理

  • Interoperability with OpenGL and Direct3D.

不过，这边应该就不会列举了～个别的说明，请参考《CUDA Programming Guide 1.0》这份文件。而这边，就只大概介绍一些 Heresy 觉得比较会用到的部分了。

记忆体管理

要说最重要、一定会用到的部分，应该就是记忆体管理的部分了！因为在 CUDA 中，要在 device 的程式中存取的资料，一定要先存在 device 上，所以在记忆体管理的部分，CUDA 提供了很类似 C 的一些函式：cudaMalloc()、cudaFree()。原则上，cudaMalloc 大致上就是 C 语言中的 malloc，也就像是 C++ 的 new；而 cudaFree 则就相当於 C 的 free，以及 C++ 的 delete。

而要使用的方法，则是要先宣告 pointer，在 allocate 记忆体给他；下面就是一个简单的范例：

float   *dev_Array;cudaMalloc( (void**)&dev_Array, ArraySize * sizeof( float ) );

这样，就可以在 device 上配置一块大小是 ArraySize 的 float 阵列了～

如果要把已经存在一般程式中的资料复制到 device 上呢？这时候就要用对应到 C 语言中 memcpy 的函式－cudaMemcpy 了～以 CUDA SDK 范例里的写法，会是：

// define, allocate on hostfloat   *host_Array = new float[100];for( int i = 0; i < 100; ++ i )        host_Array = i;// define, allocate on devicefloat   *dev_Array;cudaMalloc( (void**)&dev_Array, 100 * sizeof( float ) );// copy from host to devicecudaMemcpy>( dev_Array, host_Array, 100 * sizeof( float ),        cudaMemcpyHostToDevice );

cudaMemcpy 这个函式的四个参数依序为：目标、来源、大小、方法。最后一项方法的型别是 CUDA 的一个列举型别 cudaMemcpyKind，他有四种值：cudaMemcpyHostToHost, cudaMemcpyHostToDevice, cudaMemcpyDeviceToHost, cudaMemcpyDeviceToDevice；相当直觉的，就是代表由 host/device 复制到 device/host。所以，如果是要反过来把 device 上的资料复制回 host 的话，就只要把最后一项参数改成 cudaMemcpyDeviceToHost 就可以了！

而要释放掉 allocate 出来的记忆体空间的话，就直接使用 cudaFree 就可以了。

cudaFree( dev_Array );

实际上 CUDA 的记忆体管理除了这边提到的基本类型，还有 cudaMallocPitch(), cudaMemset(), cudaMemcpy2D(), cudaMemcpyToArray(), cudaMemcpyToSymbol() 等其他的记忆体管理函式，在这边就不一一提出来了。

 

装置管理

一般情况下，可能用不到这一部分的 API 吧～但是如果有两个以上的 CUDA Device 的话，就须要透过这些函式来选择要用的装置了～此外，也可以先透过这些函式，来确认机器上的一些基本规格。

这部分的函式不多，有下面几个：

• cudaGetDeviceCount( int* count )

  取得目前机器上，支援 CUDA 的 device 数目，会把结果存在 count 中。函式则会 return 一个 cudaError_t，用来表示 CUDA 的错误。

• cudaGetDeviceProperties( cudaDeviceProp* prop, int dev )

  取得目前机器上第 dev 个 device 的属性；prop 的结构会是：

  struct cudaDeviceProp {  char name[256];               //装置名称  size_t totalGlobalMem;        //Device 的 global memory 总量  size_t sharedMemPerBlock;     //Device 的 shared memory 总量   int regsPerBlock;     //每一个 block 的 registers 数量  int warpSize;         //warp size  size_t memPitch;              //最大的 memory pitch  int maxThreadsPerBlock;       //Block 的最大 thread 数  int maxThreadsDim[3]; //Block 中 thread 最大的维度  int maxGridSize[3];   //Grid 中 block 的最大维度  size_t totalConstMem; //Device 的 constant memory 总量  int major;            //主要版本编号  int minor;            //次要版本编号  int clockRate;                //时脉  size_t textureAlignment;      //alignment requirement mentioned};

• cudaChooseDevice( int* dev, const cudaDeviceProp* prop)

  在目前的 device 中，找一个最符合给定的 prop 的，并把他的索引直存在 dev 里。

• cudaSetDevice(int dev)

  指定现在要使用第 dev 个 device。

• cudaGetDevice(int* dev)

  取得目前正在使用的 device 的索引编号，储存在 dev 中。

 

快速数学计算

从上面的列表可以发现，在 common component 和 device component 中，都有数学函式的功能。其中，common component 中的应该算是一般的数学计算函式，而 device component 所提供的，则是精确度较低，但是速度较快的版本！

也就是说 CUDA 在 device component 的部分，提供了部分数学计算的函式，是以 GPU 来做计算的快速版本。这些函式在 CUDA 中统一以「__」来做开头，例如 __sinf(x)、__fadd_rz(x,y)、__logf(x) 等等。这些函式在《CUDA Programming Guide 1.0》的 Appendix B 中，可以找到比较完整的列表可以参考。

而在 nvcc 的编译参数中，也还提供了一个「-use_fast_math」的参数，可以强制所有可以的数学计算函数，都使用叫快速的版本来计算。这在精确度比较不重要的情况下，应该可以对效能有些增益。

---

上面讲的部分，都是 runtime API 的部分；大部分的函式，都是以 cuda 来做为开头。而实际上，CUDA 还有提供 Driver API；这一部分的函式会是以 cu 来做为开头。而许多函式，都可以和 runtime API 来做一一对应。而以文件里的说法，Driver API 应该是比 runtime API 来的低阶，但是实际的差异，Heresy 也不是很清楚就是了。

---

本文主要参考nVidia 的《CUDA Programming Guide 1.0》（PDF 文件）的第四章。