NVIDIA ® Tesla ® P4 采用革命性的 NVIDIA Pascal™ 架构，8G支持10多路720P解码。
A卡指的是ATI，一个显卡厂商，但ATI被AMD收购之后，我们也把A卡称之为AMD显卡；N卡指的是NVIDIA（英伟达），另一个显卡厂商。N卡的GPU中每个流处理器都具有完整的ALU(算术逻辑单元)功能，在发出一条操作指令时每个流处理器都能充分工作。
而A卡的GPU中每个流处理器的5个流处理单元都是固定的，不能拆开重组。
最近cuda编程在AI行业应用较多，比较热。接触的两个应用。
应用一：
利用GPU进行视频编解码，比CPU要快很多。
购买Nvidia的显卡，然后安装在Dell服务器上，linux安装相应nivida驱动后，程序可以使用nvidia给出的视频编解码库进行视频编解码，不用学习cuda编程。
应用二：
人脸识别也使用到了GPU的并行计算能力。
__global__ ”。核函数在GPU上运行。cuda语言的编译为nvcc。
CPU与GPU间内存需要显示拷贝，经过PCI总线进行交互，这里可能会有瓶颈，尤其对多路视频并发。如果使用PCI-EX16标准，数据传输率为4.8GB/S。
cuda核心是任务分解。
启动核函数：
<<<>>>运算符对kernel函数完整的执行配置参数形式是<<<Dg, Db, Ns, S>>>
参数Dg用于定义整个grid的维度和尺寸，即一个grid有多少个block。为dim3类型。Dim3 Dg(Dg.x, Dg.y, 1)表示grid中每行有Dg.x个block，每列有Dg.y个block，第三维恒为1(目前一个核函数只有一个grid)。整个grid中共有Dg.x*Dg.y个block，其中Dg.x和Dg.y最大值为65535。
参数Db用于定义一个block的维度和尺寸，即一个block有多少个thread。为dim3类型。Dim3 Db(Db.x, Db.y, Db.z)表示整个block中每行有Db.x个thread，每列有Db.y个thread，高度为Db.z。Db.x和Db.y最大值为512，Db.z最大值为62。 一个block中共有Db.x*Db.y*Db.z个thread。计算能力为1.0,1.1的硬件该乘积的最大值为768，计算能力为1.2,1.3的硬件支持的最大值为1024。
参数Ns是一个可选参数，用于设置每个block除了静态分配的shared Memory以外，最多能动态分配的shared memory大小，单位为byte。不需要动态分配时该值为0或省略不写。
参数S是一个cudaStream_t类型的可选参数，初始值为零，表示该核函数处在哪个流之中。
一般只需使用前两个参数即可。
核函数需要计算线程ID，需要根据三维grid和三维block进行计算。这里不一一列举，参考：
1. /**************************************************************/
2. // ！！！！！！！！！！！！！！注意！！！！！！！！！！！！！！！！
3. /**************************************************************/
4. // grid划分成a维，block划分成b维，
5. // 等价于
6. // blocks是a维的，Threads是b维的。
7. // 这里，本人用的是第一中说法。
8. /**************************************************************/
9.
10.
11. // 情况1：grid划分成1维，block划分为1维。
12. __device__ int getGlobalIdx_1D_1D() {
13. int threadId = blockIdx.x *blockDim.x + threadIdx.x;
14. return threadId;
15. }
16.
17. // 情况2：grid划分成1维，block划分为2维。
18. __device__ int getGlobalIdx_1D_2D() {
19. int threadId = blockIdx.x * blockDim.x * blockDim.y
20. + threadIdx.y * blockDim.x + threadIdx.x;
21. return threadId;
22. }
23.
24. // 情况3：grid划分成1维，block划分为3维。
25. __device__ int getGlobalIdx_1D_3D() {
26. int threadId = blockIdx.x * blockDim.x * blockDim.y * blockDim.z
27. + threadIdx.z * blockDim.y * blockDim.x
28. + threadIdx.y * blockDim.x + threadIdx.x;
29. return threadId;
30. }
31.
32. // 情况4：grid划分成2维，block划分为1维。
33. __device__ int getGlobalIdx_2D_1D() {
34. int blockId = blockIdx.y * gridDim.x + blockIdx.x;
35. int threadId = blockId * blockDim.x + threadIdx.x;
36. return threadId;
37. }
38.
39. // 情况5：grid划分成2维，block划分为2维。
40. __device__ int getGlobalIdx_2D_2D() {
41. int blockId = blockIdx.x + blockIdx.y * gridDim.x;
42. int threadId = blockId * (blockDim.x * blockDim.y)
43. + (threadIdx.y * blockDim.x) + threadIdx.x;
44. return threadId;
45. }
46.
47. // 情况6：grid划分成2维，block划分为3维。
48. __device__ int getGlobalIdx_2D_3D() {
49. int blockId = blockIdx.x + blockIdx.y * gridDim.x;
50. int threadId = blockId * (blockDim.x * blockDim.y * blockDim.z)
51. + (threadIdx.z * (blockDim.x * blockDim.y))
52. + (threadIdx.y * blockDim.x) + threadIdx.x;
53. return threadId;
54. }
55.
56. // 情况7：grid划分成3维，block划分为1维。
57. __device__ int getGlobalIdx_3D_1D() {
58. int blockId = blockIdx.x + blockIdx.y * gridDim.x
59. + gridDim.x * gridDim.y * blockIdx.z;
60. int threadId = blockId * blockDim.x + threadIdx.x;
61. return threadId;
62. }
63.
64. // 情况8：grid划分成3维，block划分为2维。
65. __device__ int getGlobalIdx_3D_2D() {
66. int blockId = blockIdx.x + blockIdx.y * gridDim.x
67. + gridDim.x * gridDim.y * blockIdx.z;
68. int threadId = blockId * (blockDim.x * blockDim.y)
69. + (threadIdx.y * blockDim.x) + threadIdx.x;
70. return threadId;
71. }
72.
73. // 情况9：grid划分成3维，block划分为3维。
74. __device__ int getGlobalIdx_3D_3D() {
75. int blockId = blockIdx.x + blockIdx.y * gridDim.x
76. + gridDim.x * gridDim.y * blockIdx.z;
77. int threadId = blockId * (blockDim.x * blockDim.y * blockDim.z)
78. + (threadIdx.z * (blockDim.x * blockDim.y))
79. + (threadIdx.y * blockDim.x) + threadIdx.x;
80. return threadId;
81. }
