《基于GPU加速的计算机视觉编程》学习笔记（2）

笔记（1）得到的结论

当时使用的是win10系统，自己安装的显卡驱动，再下载cuda10.1进行安装，无论是使用了自定以安装还是精简（推荐）安装，都在安装驱动那一步，显示安装失败，并且下面有一大堆东西未安装（其实是和以前安装遇到的问题是一样的，就是行不通）。

那我们应该坐以待毙么，那是不可能的，所以我们选择曲线救国，使用ubuntu16.04安装cuda。

CUDA的开发环境

拥有Ubuntu16.04系统

这里就不细讲我怎么在win10上安装的ubuntu16.04双系统了，一个原因是确实没法截图，另一个原因是双系统的安装也不是我一个人完成的，遇到了一些问题，都是同学帮我解决的，所以没有资格在这里讲解（虽然网上也没有特别详细的教程，尤其是分盘那一步）

不过，这里可以说一下在安装过程中让我印象很深的一件事吧。

就是我安装完了ubuntu16.04双系统之后，开机选择进入ubuntu，进入ubuntu的过程十分缓慢。之后的关机，reboot，也会卡住，一直卡在一个画面上只能重启。

解决方法：安装显卡驱动，安装好了，上面的问题就解决了，根本不是什么网上说的输入法，语言设置的问题。

linux下查看显卡信息

1.Ctrl+Alt+t 打开终端。
2.输入 sudo lshw -C video

我们可以看到，这里显示的信息是，GP107M（GeForce GTX 1050 Mobile）

linux下安装CUDA工具包

在 https://developer.nvidia.com/cuda-downloads 里选择对应的选项，下载安装程序。

使用ubuntu自带的apt-get，在终端里输入命令即可

sudo apt-get install nvidia-cuda-tootik

nvcc 将分别编译 .cu文件（后续测试时会讲到）中的Host和Device代码，前者是通过系统自带的GCC之类的Host代码编译器进行的，而后者则是通过CUDA C前端等一系列工具进行的。

你可以通过如下命令安装NSight Eclipse Edition（也叫做NV）用作linux下开发CUDA程序的图形化IDE环境。

sudo apt install nvidia-nsight

一个基本的CUDA C程序

hello cuda！代码如下

#include <iostream>//书中只给了这一个库函数，但是我运行失败了，因为有printf。
#include <stdio.h>//加上stdio.h就运行成功了。我是在终端上编译运行的，书上是在NV上编译运行的，难道这还会有差别？我认为可能是作者写错了？
__global__ void myfirstkernel(void){
}
int main(void){
  myfirstkernel << <1, 1 >> >();
  printf("Hello , CUDA!\n");
  return 0;
}

保存为 filename.cu文件。

nvcc -o filename filename.cu

使用nvcc的 -o命令，编译生成可执行文件

./filename

运行可执行文件，可以在终端看到

运行成功。

对如上CUDA C程序解释

global 前缀

__global__前缀是CUDA C在标准C中添加的一个限定符，它告诉编译器在这个限定符后面的函数定义应该该在设备上而不是主机上。也就是上面的函数myfirstkernel将运行在设备上。（虽然现在是空的）

main函数运行在哪里呢？

nvcc编译器将main提供给C编译器，因为它没有被global修饰，因此main函数将在主机上运行。

尖括号和数值的意思

尖括号是CUDA C的一个技巧：从主机代码调用设备代码。被称为内核调用。尖括号内的值代表我们希望在运行时从主机传递给设备的参数。比如，上面代码。<< <1,1> >> 表示，myfirstkernel函数将运行在设备上的一个块和一个线程上。

回看整个函数

myfirstkernel函数将运行在一个只有一个块和一个线程的设备上。他们通过一个成为内核启动的方法从main函数内部的主机代码启动。

使用CUDA C进行并行编程

##CUDA C中的双变量加法程序

#include <iostream>
#include <stdio.h>
#include <cuda.h>
#include <cuda_runtime.h>
//内核函数
__global__ void gpuAdd(int d_a, int d_b ,int *d_c){
  *d_c = d_a + d_b;
}

//主函数
int main(void){
  int h_c;
  int *d_c;
  cudaMalloc((void**)&d_c,sizeof(int));
  //注意这里，对于尖括号的使用，这与空格有关，不能随便改变空格的位置，一定是<<空< 参数1,参数2>>空>
  gpuAdd<< <1,1 >> >(1,4,d_c);
  cudaMemcpy(&h_c,d_c,sizeof(int),cudaMemcpyDeviceToHost);
  printf("1 + 4 = n",h_c);
  cudaFree(d_c);
  return 0;
}

其实代码很好理解，无非是使用了CUDA自带的API实现动态分配内存和释放。剩下的关键点就和Hello CUDA是一样的。

配置内核参数

为了在设备上并行启动多个程序，我们必须在内核调用中配置参数，内核调用是在内核启动配置中编写的。它们指定了Grid中块的数量，和每个块中线程的数量。我们可以并行启动很多块，而每个块又有很多个线程。通常，每个块有512或1024个线程。每个块在流多处理器上运行，一个块只能哦个的线程可以通过共享内存彼此通信。

假设我们要启动500个线程，我们可以：
1.启动一个包含500个线程的块
2.启动一个线程的500个块
3.或者两个线程，每个线程250个块
但每个块的线程数量不能超过GPU设备所支持的最大限度。

我们的目标是计算机视觉的应用程序，需要处理二维和三维图像，那么使用多维的线程和块则可以更好地进行和处理可视化。

例如：GPU支持三维网格块和三维线程块

mykernel<< <Nbx,Nby,Nbz), dim3(Ntx,Nty,Ntz) > >>()

这里的Nb表示网格中的x,y,z方向，Nt表示一个块中沿x,y,z的线程数。例如我们想要处理一个图像，可以启动一个16×16的块网络，所有的块都包含16×16的线程。语法如下：

mykernel<< <dim3(16,16),dim3(16,16)> >>()

CUDA API函数

我们会经常的在CUDA C程序中看到如下关键字， _ global _、cudaMalloc、cudaMemcpy、cudaFree。其中，global在前问中已经讲过，如果对于理解程序代码来说，不需要知道的太多，所以不再讲述。

对cudaMalloc和cudaFree，用法和malloc和free的用法是相同的，只不过是用在设备块中而已。

这里我们重点说一下cudaMemcpy。

cudaMemcpy(void * dst_ptr ,const void * src_ptr ,size_t size,enum cudaMemcpyKind kind)

example :cudaMemcpy(&h_c,d_c,sizeof(int),cudaMemcpyDeviceToHost)

这个函数类似于memcpy函数，除了是在设备块中这一点区别以外，它多了一个第四个参数：方向，即数据复制的方向：可以从主机到设备，设备到设备，主机到主机，或设备到主机。但是一定要注意，这里的方向和你设置的第一和第二个参数要是对应的。例子中，选择设备指针d_c作为源指针，h_c为目标指针，第四个参数告诉我们，是从设备显存复制到了主机内存上。

12月26日的更新到这里结束了，12月25日的内容有部分的改动。

为啥继续往下写，是因为发现自己的nvcc不能够执行global里的函数，然后调查了一圈原因也没找到，感觉是不是cuda的安装出了问题，然后又重新安装了一遍cuda，结果安装崩了，导致nvcc都用不了了。。。。。。。淦。