CUDA项目技术报告 RGB2Grey 山东科技大学并行程序设计

Technical Record&Report for CUDA_RGB2Grey Project

Copyright© workgroup41 All Rights Reserved.
Authored by ZLT - Science and Technology of Intelligence

交流使用，禁止转载

〇. 硬件平台搭建

测试平台使用Nvidia Jetson nano B02 4GB版本开发板，选择原因如下：

1. 带有GPU，满足project条件；
2. 功耗低，最大功耗模式TDP仅有10w，散热需求低，不必担心噪音问题；
3. 便携，方便跑路

缺点也有不少，比如：

1. 性能差，但是毕竟边缘计算设备，而且只卖100$，还要啥自行车；
2. 原生只支持TF卡作为存储器，IO性能低（但是可以魔改，挂载nvme SSD）；
3. 这只arm处理器挑食

就是按照Nvidia官方教程进行系统烧录，按步骤走完，鉴于手头没有多余的卡，用了一张32GB的普通闪迪作为代替，整个过程由于是成熟产品因此非常顺利，直接进大眼睛背景的Ubuntu桌面。

一. 开发环境搭建

前置芝士

• OpenCV
• NVCC
• CMake

实践过程

首先，此项目的核心需求：cuda和opencv
首先由于开发板自带nvcc开发环境，故不需要单独安装，直接使用指令查看：

nvcc -V

即可看到当前可用版本。

在此项目中opencv的作用仅限于打开及存储图片，因此选择直接安装最熟悉的opencv3.4.3
安装过程：

1. 解压

    unzip opencv-3.4.3.zip
    unzip opencv_contrib-3.4.3.zip
   

2. 生成makefile，新建build文件夹准备编译：

    cd opencv-3.4.3
    mkdir build
    cd build
   

3. 执行cmake：

   cmake -DCMAKE_INSTALL_PREFIX=/usr/local \
         -DCMAKE_BUILD_TYPE=Release \
         -DOPENCV_EXTRA_MODULES_PATH=../../opencv_contrib-3.4.3/modules \
         -DCUDA_TOOLKIT_ROOT_DIR=/usr/local/cuda-10.2 \
         -DCUDA_ARCH_BIN=5.3 \
         -DCUDA_ARCH_PTX="" \
         -DBUILD_PNG=OFF \
         -DBUILD_TIFF=OFF \
         -DBUILD_TBB=OFF \
         -DBUILD_JPEG=OFF \
         -DBUILD_JASPER=OFF \
         -DBUILD_ZLIB=OFF \
         -DBUILD_EXAMPLES=ON \
         -DBUILD_opencv_java=OFF \
         -DBUILD_opencv_python2=OFF \
         -DBUILD_opencv_python3=ON \
         -DENABLE_PRECOMPILED_HEADERS=OFF \
         -DWITH_OPENCL=OFF \
         -DWITH_OPENMP=OFF \
         -DWITH_FFMPEG=ON \
         -DWITH_GSTREAMER=OFF \
         -DWITH_GSTREAMER_0_10=OFF \
         -DWITH_CUDA=ON \
         -DWITH_GTK=ON \
         -DWITH_VTK=OFF \
         -DWITH_TBB=ON \
         -DWITH_1394=OFF \
         -DWITH_OPENEXR=OFF \
         -DINSTALL_C_EXAMPLES=ON \
         -DINSTALL_TESTS=OFF \
         ..
   
   

4. 编译

    make -j4
   

使用全部4个核心进行编译，但即使如此还是很慢
特别的，在这台开发板上，编译会出很多问题：

• fatal error: boostdesc_bgm.i: No such file or directory
  这是由于其中几个库被墙无法成功下载，解决方法是在网上找到资源手动添加到相应目录下；
• fatal error: opencv2/xfeatures2d/cuda.hpp: No such file or directory
  这类错误在进度70%后就会频繁出现，特别是到了97%甚至99%的时候（没事儿，心态很棒，手动微笑）
  解决办法也很暴力，错误原因是出错的代码中找不到文件，只需要把include里的相对路径改成绝对路径就行了，至于具体要改哪个代码，改成什么路径，抛出的错误中会告诉我们。

• 这一步是安装过程中最难的，需要解决上述问题的无数次重复出现，但还好最后成功了

5. 安装

    sudo make install
   

6. 这时就可以高兴的查看安装情况了！

   opencv版本查看：

    pkg-config --modversion opencv
   

二、交叉编译

前置芝士：

• 静态库
• //动态库
• CMake

经过研（mo）究（yu）发现，实现交叉编译的方法很多，但总体思想是将.cu文件编译出的东西，作为最终编译出的可执行程序的链接依赖。
我们使用gcc/g++编译c/c++部分代码，nvcc编译cuda代码部分
将cuda部分写成接口:
void API(){…}
在c/c++代码中通过添加接口声明：
extend “C” void API（）{…}
在了解了静态库和动态库后，发现可以使用cmake来将这些命令制作成CMakeList来方便的编译；将cuda部分作为一个项目并制作为库，由主项目调用。
这里选择了将cuda项目生成静态库，写在cuda的CMakeLists.txt中。
CMakeLists中的编译选项使用了文末链接中的写法，具体原理参考CMake手册
目录树：

RGB2Grey
	|____ main.cpp
	|____ CMakeLists.txt
	|____ cuda
			|____ foo.cuh
			|____ foo.cu
			|____ CMakeLists.txt

三、代码实现

• main.cpp:

图像相关：读取彩色图片为三通道8bit位图（8UC3），保存该三阶张量使用cv::Mat，由于uchar类型就是8位无符号数据类型，因此这张图片在内存中是这样$(width\ast height)\ast 3\ast unsignedchar$，在cuda的库中有uchar3类型，可以表示每一个像素（整张图也就是$(width\ast height)\ast uchar3$）但在当前main之下没有调用cuda库，因此避免出现，只需要将Mat.data作为实参丢进useCUDA函数即可；

注：8bit色深，直观来说是我们在PS里选取颜色时，那个三元组，例如$(255,255,255)$表示白色，对于其中任意一元，它的取值范围是0~255共$256=2^8$个数值，也就是从黑到白分了256个档（叫做灰度），此时，需要8个bit用来存储这一个灰度值；**回到三元组来，**三个通道（RGB）每一个都是8bit，那么总共就是3*8bit，也就是uchar3类型（见下方代码）。用微积分的极限思想来理解一下，如果我们只有1bit也就是2个数值来表示，那么只能从黑到白划分0和1两个灰度，就是只有黑和白（叫做二值图）；若是2bit，那就是4个灰度，黑白之间多了深浅两个灰色，以此类推。因此，bit数越大，黑白过渡越细腻，对应的图片显示效果越棒，同时数据量也越大，这就是为什么显示器10bit的比8bit贵得多，相机能达到12bit，电影机14bit甚至更高（它贵得有理）

//uchar3定义方式
struct uchar3{
   	unsigned char x, y, z;//unsigned char为8bit无符号数据类型
}

 | 方法 |     功能|
 | :-- | --------|
 | cv::imwrite()   |   保存图片序列，第一个参数是string类型作为保存路径，第二个是内存数据（Mat类型）|
 | cv::Mat    |   读入图片保存类型，在这里我们可以用uchar*传参|

处理图像序列部分代码：

    string first_file = "/home/zlt/cvcvcv/cross_compiler_rgb2grey/img/%6d.jpg";
    VideoCapture sequence(first_file);

    Mat srcImage;
    int frameID = 0;
    while(2){
        sequence >> srcImage;
        frameID++;
        if(srcImage.empty()) break;

        const uint imgheight = srcImage.rows;
        const uint imgwidth = srcImage.cols;
        Mat grayImage(imgheight, imgwidth, CV_8UC1, Scalar(0));

        grayImage.data = useCUDA(srcImage.data, imgheight, imgwidth);
            //the type of 'srcImage.data' is 'unsigned char'
        String SaveName = "/home/zlt/cvcvcv/cross_compiler_rgb2grey/out/" + to_string(frameID) + ".jpg";
        imwrite(SaveName, grayImage);
    }
    cout << "End of Sequence" << endl;

• foo.cu中的useCUDA()函数：
  用uchar*指针，没有出错（
  后面两个是图片高度和宽度

    unsigned char* useCUDA(uchar* inn, unsigned int imgheight, unsigned int imgwidth)
  

开d_in和d_out指向显存空间，d_return指向内存空间作为返回的灰度图像（in是三阶张量，用uchar3；out是一阶张量，用uchar）

    uchar* d_out;//gpumemory
    uchar* d_return;//memory
    uchar3* d_in;

使用相应类型开存储空间

    cudaMalloc((void**)&d_in, imgheight * imgwidth * sizeof(uchar3));
    cudaMalloc((void**)&d_out, imgheight * imgwidth * sizeof(unsigned char));
    d_return = new uchar[imgheight * imgwidth];

开显存、拷显存、核函数、拷回内存，和前面的作业同理；
注意：彩色图像是三通道

• foo.cuh:
  声明前面加上 extern "C"
• 核函数：
  就是个矩阵数乘而已，比作业里的矩阵相乘还简单

四、运行效果

我们顺手使用MOT-20数据集作为测试图像，图像大小1920*1080，选取了其中的150帧进行输入，如期得到了结果。
使用nvprof进行分析：

输出图像序列：

参考资料

[1] Jetson Nano 从头配置OpenCV+CUDA+QT完整流程
[2] nvcc gcc g++混合编译器编程
[3] CUDA编译（二）—用CMake混合编译C++与cuda