闵大荒之旅(五) ----- OpenCV与CUDA编程的结合

时间:2024-02-19 13:03:17

参考http://www.cnblogs.com/dwdxdy/p/3528711.html博客,加之以实践环节,我们可以知道有几种使用到GPU运算的方法:

                    1.利用OpenCV中提供的GPU模块

                2.单独使用Cuda API编程

                3.利用OpenCV中提供接口,并结合Cuda API编程

如果仅仅使用OpenCV中的GPU函数,就像博客(三)中演示一下,的确非常的简单而且可以得到比较理想的效果,但是缺点也是显然的,这种直接利用别人的函数是非常不灵活的。很多情况下,并行计算都会存在一个最优的问题,如何使得实际问题计算最优化,这就需要我们自己去思考,而不是仅仅调用别人的函数;

如果光单独使用CUDA编程,那这个工作量是可想而知的了,上文中有说道,GPU是处理计算量大计算繁琐复杂的部分,而其他的部分交给CPU去完成就是,那么这一部分完全可以使用直接提供的函数,例如OpenCV,所以,光使用CUDA进行编程,我也觉得没有必要;

最后,针对我所研究的课题,OpenCV结合CUDA编程是再适合不过的了。

 

这一次,我主要介绍在Ubuntu下如何将OpenCV与CUDA结合起来,也就是如何将.CU文件和.CPP文件结合起来得到想要的结果。这一部分我在网上查阅过大量的资料,但是资料非常的少,有的也只是很含糊的一笔带过,虽然方法非常简单,但是这里我还是详细地说明一下。

 

首先,我们的目的是:OpenCV+CUDA,完成对图像的灰度化

方法非常的简单,就是利用平均值算法 gray = ( r + g + b ) / 3得到对应像素点的值,那么这个计算部分我们就交给CUDA完成,最后结果我们将返还给CPU进行进一步处理以及结果显示。

先上代码:

   test.cu文件

#include <opencv2/core/cuda_devptrs.hpp>
#include "device_launch_parameters.h"
#include "cuda.h"
#include "cuda_runtime_api.h"
#include "book.h"
using namespace cv;
using namespace cv::gpu;
//自定义内核函数
__global__ void swap_rb_kernel(const PtrStepSz<uchar3> src, PtrStep<unsigned char> dst)
{
  int y = blockIdx.y*blockDim.y+threadIdx.y;
  int x = blockIdx.x*blockDim.x+threadIdx.x;  
  if(x < src.cols && y < src.rows) 
  {
       dst(y, x) = (src(y, x).x + src(y, x).y + src(y, x).z) / 3;
  }
}

extern "C" void swap_rb_caller(const PtrStepSz<uchar3>& src, PtrStep<unsigned char> dst, cudaStream_t stream)
{
  int uint = 8;

  dim3 grid(src.cols + uint-1/uint,src.rows + uint-1/uint);
  dim3 block(uint,uint);
  swap_rb_kernel<<<grid,block>>>(src, dst);
  if(stream == 0)
        cudaDeviceSynchronize();
}

  testcpp.cpp

#include "stdio.h"
#include "opencv2/core/core.hpp"
#include "opencv2/highgui/highgui.hpp"
#include "opencv2/imgproc/imgproc.hpp"
#include "opencv2/gpu/gpu.hpp"
#include "opencv2/ml/ml.hpp"
#include "opencv2/gpu/gpu.hpp"
#include <opencv2/gpu/stream_accessor.hpp>
#include "cuda.h"
#include "cuda_runtime_api.h"
#include <cmath>
#include <iostream>
#ifdef _DEBUG
#pragma comment(lib, "opencv_core249d.lib")
#pragma comment(lib, "opencv_imgproc249d.lib")
#pragma comment(lib, "opencv_highgui249d.lib")
#else
#pragma comment(lib, "opencv_core249.lib")
#pragma comment(lib, "opencv_imgproc249.lib")
#pragma comment(lib, "opencv_highgui249.lib")
#endif // DEBUG
using namespace cv;
using namespace cv::gpu;
extern "C" void swap_rb_caller(const PtrStepSz<uchar3>& src, PtrStep<uchar3> dst, cudaStream_t stream);
void swap_rb(const GpuMat& src, GpuMat& dst, Stream& stream = Stream::Null())
{
    CV_Assert(src.type() == CV_8UC3);
    dst.create(src.size(), CV_8UC1);
    cudaStream_t s = StreamAccessor::getStream(stream);
    swap_rb_caller(src, dst, s);
    //cuResize(src, src.getWidth(), src.getHeight(), dst, dst.getWidth(), dst.getHeight())
}

int main()
{
    cudaSetDevice(0);
    cudaFree(0);    
    printf("hello\n");
    Mat image = imread("test.png");
    imshow("src",image);
    GpuMat gpuMat,output;
    output.create(image.size(), CV_8UC1);
    Mat result;
    result.create(image.size(), CV_8UC1);
    double start = (double)getTickCount();
    gpuMat.upload(image);
    swap_rb(gpuMat,output);
    output.download(result);
    double t = ((double)getTickCount() - start)/getTickFrequency();
    printf("gpu time is : %d ms\n", (int)(1000 * t));
    Mat result2;
    start = (double)getTickCount();
    cvtColor(image, result2, COLOR_BGR2GRAY);
    t = ((double)getTickCount() - start)/getTickFrequency();
    printf("cpu time is : %d ms\n", (int)(1000 * t));
    imshow("gpu", result);
    imshow("cpu", result2);
    waitKey(0);
    return 0;
}

这么一看,这个问题实际上就变成了CPP中调用CU中定义的函数

这里我相信大家已经看到了, 从CPU传送数据到GPU实际上就是GpuMat->PtrStepSz的过程,注意,如果Mat是3通道的,就要PtrStepSz<uchar3>,一通道的则为PtrStepSz<unsigned char>。

关于PtrStepSz的具体信息,可以参考http://docs.opencv.org/2.4/modules/gpu/doc/data_structures.html

 

有了程序,接下来我们尝试运行它。

 

首先,编译.CU文件,在终端下输入命令:

nvcc test.cu -c -o test

编译成功后生成了一个test的文件,类型暂不明

 

终端输入命令:

ar cqs libtest5.a test

则生成了一个libtest5.a的静态库,这个库名可以自己另取

 

接下来我们要将库考入ubuntu默认搜素路径:

sudo cp libtest5.a /usr/lib/

 

于是,接下来可以编译cpp程序:

g++ testcpp.cpp -lopencv_core -lopencv_imgproc -lopencv_highgui -lopencv_calib3d -lopencv_contrib -lopencv_features2d -lopencv_flann -lopencv_gpu -lopencv_legacy -lopencv_ml -lopencv_objdetect -lopencv_photo -lopencv_stitching -lopencv_superres -lopencv_video -lopencv_videostab -L./ -ltest5 -o testcpp -lcudart -lpthread

这里一定要注意了,如果报出DSO missing from command line这个错误,第一看看相应的包是否导入/usr/iib/中,第二,如果已经导入成功了,就将-l语句写到命令的最后,编译才不会出错!!!!一定注意!

 

于是,就可以生成我们熟悉的可执行文件

 

运行得到结果:

 

运行成功,但是时间表明,对于灰度化处理,gpu运行时间并没有cpu快,这说明,传输消耗的时间远比运行时间长,这样使用gpu计算是划不来的。

但是本人的目的是成功运行cuda+opencv程序,至此,已经成功!