之前給大家分享過opencv在jetson nano 2gb和ubuntu設備中使用并且展示了一些人臉識別等的小demo。但是對于圖像處理，使用gpu加速是很常見 .(以下概念介紹內(nèi)容來自百科和網(wǎng)絡其他博主文章)

GPU介紹(從GPU誕生之日起，GPU的設計邏輯與CPU的設計邏輯相差很多。GPU從誕生之日起，它的定位是3D圖形渲染設備。在設計GPU時從其功能出發(fā)，把更多的晶體管用于數(shù)據(jù)處理。這使得GPU相比CPU有更強的單精度浮點運算能力。人們?yōu)榱顺浞掷肎PU的性能，使用了很多方法。這)加速處理是比較常見的。

在而GPU加速的軟件實現(xiàn)中我們可能會聽到 opengl opencl cuda這些名詞。下面再給大家分別介紹一下這幾種區(qū)別：

OpenGL（英語：Open Graphics Library，譯名：開放圖形庫或者“開放式圖形庫”）是用于渲染2D、3D矢量圖形的跨語言、跨平臺的應用程序編程接口（API）。這個接口由近350個不同的函數(shù)調(diào)用組成，用來繪制從簡單的圖形比特到復雜的三維景象。而另一種程序接口系統(tǒng)是僅用于Microsoft Windows上的Direct3D。OpenGL常用于CAD、虛擬現(xiàn)實、科學可視化程序和電子游戲開發(fā)。

OpenCl(是由蘋果（Apple）公司發(fā)起，業(yè)界眾多著名廠商共同制作的面向異構(gòu)系統(tǒng)通用目的并行編程的開放式、免費標準，也是一個統(tǒng)一的編程環(huán)境。便于軟件開發(fā)人員為高性能計算服務器、桌面計算系統(tǒng)、手持設備編寫高效輕便的代碼，而且廣泛適用于多核心處理器(CPU)、圖形處理器(GPU)、Cell類型架構(gòu)以及數(shù)字信號處理器(DSP)等其他并行處理器，在游戲、娛樂、科研、醫(yī)療等各種領域都有廣闊的發(fā)展前景。)

從2007年以后，基于CUDA和OpenCL這些被設計成具有近似于高階語言的語法特性的新GPGPU語言，降低了人們使用GPGPU的難度，平緩了開始時的學習曲線。使得在GPGPU領域，OpenGL中的GLSL逐漸退出了人們的視線。來源：

簡單說OpenCL與OpenGL一樣，都是基于硬件API的編程。OpenGL是針對圖形的，而OpenCL則是針對并行計算的API，而針對并行計算下面還有一個cuda。

CUDA（Compute Unified Device Architecture，統(tǒng)一計算架構(gòu)）是由英偉達NVIDIA所推出的一種整合技術，是該公司對于GPGPU的正式名稱。透過這個技術，使用者可利用NVIDIA的GeForce 8以后的GPU和較新的Quadro GPU進行計算。亦是首次可以利用GPU作為C-編譯器的開發(fā)環(huán)境。NVIDIA行銷的時候，往往將編譯器與架構(gòu)混合推廣，造成混亂。實際上，CUDA可以相容OpenCL或者自家的C-編譯器。無論是CUDA C-語言或是OpenCL，指令最終都會被驅(qū)動程式轉(zhuǎn)換成PTX代碼，交由顯示核心計算。

通俗的介紹，cuda是nvidia公司的生態(tài)，它前者是配備完整工具包、針對單一供應商(NVIDIA)的成熟的開發(fā)平臺，opencl是一個開源的標準。

CUDA是NVIDIA GPU編程語言， OpenCL是異構(gòu)計算庫。CUDA和C++雖然都可以用nvcc編譯，但C++只能在CPU上跑，CUDA只能在GPU上跑；而OpenCL并不局限于某個計算設備，旨在將同樣的任務通過其提供的抽象接口在多種硬件上運行（CPU，GPU，FPGA，etc）

跨平臺性和通用性上OpenCL占有很大優(yōu)勢（這也是很多National Laboratory使用OpenCL進行科學計算的最主要原因）。OpenCL支持包括ATI,NVIDIA,Intel,ARM在內(nèi)的多類處理器，并能支持運行在CPU的并行代碼，同時還獨有Task-Parallel Execution Mode，能夠更好的支持Heterogeneous Computing。這一點是僅僅支持數(shù)據(jù)級并行并僅能在NVIDIA眾核處理器上運行的CUDA無法做到的。

在開發(fā)者友好程度CUDA在這方面顯然受更多開發(fā)者青睞。原因在于其統(tǒng)一的開發(fā)套件(CUDA Toolkit, NVIDIA GPU Computing SDK以及NSight等等)、非常豐富的庫(cuFFT, cuBLAS, cuSPARSE, cuRAND, NPP, Thrust)以及NVCC(NVIDIA的CUDA編譯器)所具備的PTX(一種SSA中間表示，為不同的NVIDIA GPU設備提供一套統(tǒng)一的靜態(tài)ISA)代碼生成、離線編譯等更成熟的編譯器特性。相比之下，使用OpenCL進行開發(fā)，只有AMD對OpenCL的驅(qū)動相對成熟。?來源

今天的主題來自介紹cuda的使用，前面這部分概念性的介紹（來源都是網(wǎng)絡，只有一小部分是自己寫的）只是幫助大家好理解今天要使用的cuda，那么開始進入正題。（自己本身的電腦帶NVIDIA的顯卡，以及手里面有jetson nano的NVIDIA板卡，所以才使用cuda，其他朋友使用請注意自己手中的硬件是否有NVIDIA的顯卡）。

接下來，我大致分為三個部分介紹，一、cuda并行計算原理介紹 ，二、cuda環(huán)境安裝和直接使用，三、opencv中cuda安裝和使用

歡迎關注微信公眾號：羽林君，或者添加作者個人微信：become_me

cuda并行計算原理介紹

GPU背景介紹

GPU里有很多Compute Unit(計算單元), 這些單元是由專門的處理邏輯, 很多register, 和L1 cache 組成. Memory access subsystem 把GPU 和RAM 連起來(通常是個L2 cache). Threads 以SIMT的形式運行: 多個thread共享一個instruction unit. 對NV GPU來說, 32 thread 組成一個warps, 對AMD GPU 來說, 64 threads 叫做一個wave fronts. 本文只使用warp的定義. 想達到最高速度, 一定要討論SIMT, 因為它影響著memory access 也會造成code的序列化(serialization, parallelism的反面)

kernel function里會指出哪些code由一條thread處理, host program會決定多少條thread來處理一個kernel. 一個work group 里的thread可以通過barrier synchronization, 共享L1 cache 來互相協(xié)力. 再由compute unit處理這些work group. 能同時被conpute unit 運行的work group數(shù)量有限, 所以很多得等到其他完成之后再運行. 需要處理的work group的size和數(shù)量還有最大并行數(shù)量被稱為kernel的execution configuration(運行配置).

kernel 的占用率就是指同時運行的thread數(shù)量除以最大數(shù)量. 傳統(tǒng)建議就是提升這個占用率來獲得更好的性能, 但也有一些其他因素比如ILP, MLP, instruction latencies, 也很關鍵.內(nèi)容來源

GPU架構(gòu)特點

首先我們先談一談串行計算和并行計算。我們知道，高性能計算的關鍵利用多核處理器進行并行計算。

當我們求解一個計算機程序任務時，我們很自然的想法就是將該任務分解成一系列小任務，把這些小任務一一完成。在串行計算時，我們的想法就是讓我們的處理器每次處理一個計算任務，處理完一個計算任務后再計算下一個任務，直到所有小任務都完成了，那么這個大的程序任務也就完成了。

串行計算的缺點非常明顯，如果我們擁有多核處理器，我們可以利用多核處理器同時處理多個任務時，而且這些小任務并沒有關聯(lián)關系（不需要相互依賴，比如我的計算任務不需要用到你的計算結(jié)果），那我們?yōu)槭裁催€要使用串行編程呢？為了進一步加快大任務的計算速度，我們可以把一些獨立的模塊分配到不同的處理器上進行同時計算（這就是并行），最后再將這些結(jié)果進行整合，完成一次任務計算。下圖就是將一個大的計算任務分解為小任務，然后將獨立的小任務分配到不同處理器進行并行計算，最后再通過串行程序把結(jié)果匯總完成這次的總的計算任務。

所以，一個程序可不可以進行并行計算，關鍵就在于我們要分析出該程序可以拆分出哪幾個執(zhí)行模塊，這些執(zhí)行模塊哪些是獨立的，哪些又是強依賴強耦合的，獨立的模塊我們可以試著設計并行計算，充分利用多核處理器的優(yōu)勢進一步加速我們的計算任務，強耦合模塊我們就使用串行編程，利用串行+并行的編程思路完成一次高性能計算。

接下來我們談談CPU和GPU有什么區(qū)別，他們倆各自有什么特點，我們在談并行、串行計算時多次談到“多核”的概念，現(xiàn)在我們先從“核”的角度開始這個話題。首先CPU是專為順序串行處理而優(yōu)化的幾個核心組成。而GPU則由數(shù)以千計的更小、更高效的核心組成，這些核心專門為同時處理多任務而設計，可高效地處理并行任務。也就是，CPU雖然每個核心自身能力極強，處理任務上非常強悍，無奈他核心少，在并行計算上表現(xiàn)不佳；反觀GPU，雖然他的每個核心的計算能力不算強，但他勝在核心非常多，可以同時處理多個計算任務，在并行計算的支持上做得很好。

GPU和CPU的不同硬件特點決定了他們的應用場景，CPU是計算機的運算和控制的核心，GPU主要用作圖形圖像處理。圖像在計算機呈現(xiàn)的形式就是矩陣，我們對圖像的處理其實就是操作各種矩陣進行計算，而很多矩陣的運算其實可以做并行化，這使得圖像處理可以做得很快，因此GPU在圖形圖像領域也有了大展拳腳的機會。下圖表示的就是一個多GPU計算機硬件系統(tǒng)，可以看出，一個GPU內(nèi)存就有很多個SP和各類內(nèi)存，這些硬件都是GPU進行高效并行計算的基礎。

現(xiàn)在再從數(shù)據(jù)處理的角度來對比CPU和GPU的特點。CPU需要很強的通用性來處理各種不同的數(shù)據(jù)類型，比如整型、浮點數(shù)等，同時它又必須擅長處理邏輯判斷所導致的大量分支跳轉(zhuǎn)和中斷處理，所以CPU其實就是一個能力很強的伙計，他能把很多事處理得妥妥當當，當然啦我們需要給他很多資源供他使用（各種硬件），這也導致了CPU不可能有太多核心（核心總數(shù)不超過16）。而GPU面對的則是類型高度統(tǒng)一的、相互無依賴的大規(guī)模數(shù)據(jù)和不需要被打斷的純凈的計算環(huán)境，GPU有非常多核心（費米架構(gòu)就有512核），雖然其核心的能力遠沒有CPU的核心強，但是勝在多， 在處理簡單計算任務時呈現(xiàn)出“人多力量大”的優(yōu)勢，這就是并行計算的魅力。內(nèi)容來源

整理一下兩者特點就是：

• 1.CPU：擅長流程控制和邏輯處理，不規(guī)則數(shù)據(jù)結(jié)構(gòu)，不可預測存儲結(jié)構(gòu)，單線程程序，分支密集型算法2.GPU：擅長數(shù)據(jù)并行計算，規(guī)則數(shù)據(jù)結(jié)構(gòu)，可預測存儲模式

CUDA存儲器類型：
每個線程擁有自己的?register寄存器?and?loacal?memory?局部內(nèi)存
每個線程塊擁有一塊?shared?memory?共享內(nèi)存
所有線程都可以訪問?global?memory?全局內(nèi)存

還有，可以被所有線程訪問的
?????只讀存儲器：
?????constant?memory?(常量內(nèi)容)?and?texture?memory
????
a.?寄存器Register
???寄存器是GPU上的高速緩存器，其基本單元是寄存器文件，每個寄存器文件大小為32bit.
?? Kernel中的局部(簡單類型)變量第一選擇是被分配到Register中。
???????特點：每個線程私有，速度快。

b.?局部存儲器?local?memory

???當register耗盡時，數(shù)據(jù)將被存儲到local?memory。
???如果每個線程中使用了過多的寄存器，或聲明了大型結(jié)構(gòu)體或數(shù)組，
???或編譯器無法確定數(shù)組大小，線程的私有數(shù)據(jù)就會被分配到local?memory中。
???
???????特點：每個線程私有；沒有緩存，慢。
???????注：在聲明局部變量時，盡量使變量可以分配到register。如：
???????unsigned?int?mt[3];
???????改為：unsigned int mt0, mt1, mt2;

c.?共享存儲器?shared?memory
???????可以被同一block中的所有線程讀寫
???????特點：block中的線程共有；訪問共享存儲器幾乎與register一樣快.

d.?全局存儲器?global?memory
??特點：所有線程都可以訪問；沒有緩存

e.?常數(shù)存儲器constant?memory
???用于存儲訪問頻繁的只讀參數(shù)
???特點：只讀；有緩存；空間小(64KB)
???注：定義常數(shù)存儲器時，需要將其定義在所有函數(shù)之外，作用于整個文件

f.?紋理存儲器?texture?memory
???????是一種只讀存儲器，其中的數(shù)據(jù)以一維、二維或者三維數(shù)組的形式存儲在顯存中。
???在通用計算中，其適合實現(xiàn)圖像處理和查找，對大量數(shù)據(jù)的隨機訪問和非對齊訪問也有良好的加速效果。
???????特點：具有紋理緩存，只讀。
threadIdx，blockIdx,?blockDim,?gridDim之間的區(qū)別與聯(lián)系
?在啟動kernel的時候，要通過指定gridsize和blocksize才行
?dim3?gridsize(2,2);???//?2行*2列*1頁?形狀的線程格，也就是說?4個線程塊
??? gridDim.x，gridDim.y，gridDim.z相當于這個dim3的x，y，z方向的維度，這里是2*2*1。
序號從0到3，且是從上到下的順序，就是說是下面的情況:
?具體到?線程格?中每一個?線程塊的 id索引為：
?
??? grid 中的 blockidx 序號標注情況為：??????0???? 2 
??????????????????????????????????????????1?????3
???????????
???dim3?blocksize(4,4);??//?線程塊的形狀，4行*4列*1頁，一個線程塊內(nèi)部共有?16個線程
?
???blockDim.x，blockDim.y，blockDim.z相當于這個dim3的x，y，z方向的維度，
???這里是4*4*1.序號是0-15，也是從上到下的標注：

block中的?threadidx?序號標注情況???0???????4???????8??????12?
??????1???????5???????9??????13
??????2???????6???????10?????14
??????3???????7???????11?????15
1.??1維格子，1維線程塊，N個線程======

?實際的線程id?tid?=??blockidx.x?*?blockDim.x?+?threadidx.x
?塊id???0?1?2?3?
?線程id?0?1?2?3?4
2.?1維格子，2D維線程塊
?塊id???1?2?3?
?線程id??0??2
????????1??3
?????????塊id???????????塊線程總數(shù)
?實際的線程id?tid?=??blockidx.x?*?blockDim.x?*?blockDim.y?+?
?????????當前線程行數(shù)????每行線程數(shù)
??????threadidx.y?*?blockDim.x???+?
?????????當前線程列數(shù)
??????threadidx.x

更加詳細信息，大家可以參考這兩篇文章：

https://www.cnblogs.com/skyfsm/p/9673960.html

https://github.com/Ewenwan/ShiYanLou/tree/master/CUDA

安裝cuda使用環(huán)境

cuda使用中 首先我們直接安裝CUDA toolkit （cuDNN是用于配置深度學習使用，本次我沒有使用到，就是下載cuDNN的庫放置到對應的鏈接庫目錄，大家可以自行去搜索安裝），然后直接調(diào)用cuda的api進行一些計算的測試，還有是可以安裝opencv中的cuda接口，通過opencv中的cuda使用。

直接使用cuda

直接安裝CUDA ，首先需要下載安裝包

CUDA toolkit（toolkit就是指工具包）cuDNN 注：cuDNN 是用于配置深度學習使用（這次我沒有使用就沒有下載）

官方教程

CUDA：Installation Guide Windows :: CUDA Toolkit Documentation（鏈接）

注：按照自己版本去查看資料，打開鏈接后，右上角處有一個older選項，大家可以選擇自己對應的版本文檔。

cuDNN：Installation Guide :: NVIDIA Deep Learning cuDNN Documentation（鏈接）

安裝cuda要注意一個版本匹配問題，就是我們不同的顯卡是對應不同的驅(qū)動。所以第一步就是查看顯卡驅(qū)動版本，我使用的是Ubuntu，和window安裝有些區(qū)別。第一如果我們安裝過自己的顯卡驅(qū)動，使用 nvidia-smi命令查看支持的cuda版本

如果沒有安裝可以選擇ubuntu的bash界面搜索附加驅(qū)動，然后進行安裝。參考

當然由于ubuntu本身也有不同的版本支持，大家也可以直接到驅(qū)動的官網(wǎng)進行查找，里面有專門的ubuntu20適配版本選項，其中由于我使用的是ubuntu20，而CUDA11.0以上版本才支持Ubuntu20.04，所以直接就選擇了最新的。

官方地址：https://developer.nvidia.com/cuda-toolkit-archive

選擇CUDA Toolkit 11.6.0之后，里面有幾種選項，分別是本地deb安裝、網(wǎng)絡下載安裝和可執(zhí)行文件安裝，我因為網(wǎng)絡不太順暢原因選擇了第一個選項，最后下載下載下來的deb文件有2.7G，大家做好下載大文件的準備。

wget?https://developer.download.nvidia.com/compute/cuda/repos/ubuntu2004/x86_64/cuda-ubuntu2004.pin
sudo?mv?cuda-ubuntu2004.pin?/etc/apt/preferences.d/cuda-repository-pin-600
wget?https://developer.download.nvidia.com/compute/cuda/11.6.0/local_installers/cuda-repo-ubuntu2004-11-6-local_11.6.0-510.39.01-1_amd64.deb
sudo?dpkg?-i?cuda-repo-ubuntu2004-11-6-local_11.6.0-510.39.01-1_amd64.deb
sudo?apt-key?add?/var/cuda-repo-ubuntu2004-11-6-local/7fa2af80.pub
sudo?apt-get?update
sudo?apt-get?-y?install?cuda

這時候cuda就安裝好了，接下來就是使用了。

大家可以使用cuda命令，tab就可以出現(xiàn)對應的命令，或者使用nvcc編譯對應的cuda的.cu文件進行驗證。

大家可以通過以下鏈接里面的文檔指導寫符合自己需要的功能：https://docs.nvidia.com/cuda/archive/11.6.0/cuda-c-programming-guide/index.html

下面官網(wǎng)API文件目錄：

Versioned Online Documentation

官方示例：

接下來開始進行cuda的使用：輸出hello world

#include?"cuda_runtime.h"
#include?"device_launch_parameters.h"
#include?<stdio.h>
#include?<iostream>
using?namespace?std;

__global__?void?kernel(void)?{?//帶有了__global__這個標簽，表示這個函數(shù)是在GPU上運行
??printf("hello?world?gpu?n");
}
int?main()?{
? kernel<<<1, 1>>>();//調(diào)用除了常規(guī)的參數(shù)之外，還增加了<<<>>>修飾,其中第一個1，代表線程格里只有一個線程塊；第二個1，代表一個線程塊里只有一個線程。
????cudaError_t?cudaStatus;

????cudaStatus?=?cudaDeviceSynchronize();
????if?(cudaStatus?!=?cudaSuccess)?{
????????fprintf(stderr,?"cudaDeviceSynchronize?returned?error?code?%d?after?launching?addKernel!n",?cudaStatus);
????}

??return?0;
}

這里面記得使用cudaDeviceSynchronize函數(shù)，等待GPU完成運算。如果不加的話，你將看不到printf的輸出

編譯選項 nvcc test_cuda_hello.cu

這個程序中GPU調(diào)用的函數(shù)和普通的C程序函數(shù)的區(qū)別

函數(shù)的調(diào)用除了常規(guī)的參數(shù)之外，還增加了<<<>>>修飾。調(diào)用通過<<<參數(shù)1,參數(shù)2>>>，用于說明內(nèi)核函數(shù)中的線程數(shù)量，以及線程是如何組織的。以線程格（Grid）的形式組織，每個 線程格 由若干個 線程塊（block）組成，而每個線程塊又由若干個線程（thread）組成。是以block為單位執(zhí)行的。

接下來，我們再來使用一個demo，使用一個簡單的計算，將其放置于GPU進行計算

//?輸入變量?以指針方式傳遞===================================
#include?<iostream>
#include?<cuda.h>
#include?<cuda_runtime.h>
#include?<stdio.h>

//?輸入變量?全部為指針??類似數(shù)據(jù)的引用
__global__?void?gpuAdd(int?*d_a,?int?*d_b,?int?*d_c)?
{
?*d_c?=?*d_a?+?*d_b;
}

int?main(void)?
{
?//?CPU變量
?int?h_a,h_b,?h_c;
?//?CPU??指針?變量?指向?GPU數(shù)據(jù)地址
?int?*d_a,*d_b,*d_c;
?//?初始化CPU變量
?h_a?=?1;
?h_b?=?4;
??
?//?分配GPU?變量內(nèi)存
?cudaMalloc((void**)&d_a,?sizeof(int));
?cudaMalloc((void**)&d_b,?sizeof(int));
?cudaMalloc((void**)&d_c,?sizeof(int));
??
?//?輸入變量?CPU?拷貝到?GPU???右?到?左
?cudaMemcpy(d_a,?&h_a,?sizeof(int),?cudaMemcpyHostToDevice);
?cudaMemcpy(d_b,?&h_b,?sizeof(int),?cudaMemcpyHostToDevice);
??
?//?調(diào)用核函數(shù)
?gpuAdd?<<?<1,?1?>>?>?(d_a,?d_b,?d_c);
??
?//?拷貝GPU數(shù)據(jù)結(jié)果?d_c?到?CPU變量
?cudaMemcpy(&h_c,?d_c,?sizeof(int),?cudaMemcpyDeviceToHost);
?printf("Passing?Parameter?by?Reference?Output:?%d?+?%d?=?%dn",?h_a,?h_b,?h_c);
??
?//?清理GPU內(nèi)存?Free?up?memory?
?cudaFree(d_a);
?cudaFree(d_b);
?cudaFree(d_c);
?return?0;
}

CUDA代碼中比較重要的函數(shù):

cudaMalloc與C語言中的malloc函數(shù)一樣，只是此函數(shù)在GPU的內(nèi)存你分配內(nèi)存。

addKernel<<<1, size>>>

這里就涉及了GPU和主機之間的內(nèi)存交換了，cudaMalloc是在GPU的內(nèi)存里開辟一片空間，
然后通過操作之后，這個內(nèi)存里有了計算出來內(nèi)容，再通過cudaMemcpy這個函數(shù)把內(nèi)容從GPU復制出來?？梢栽谠O備代碼中使用cudaMalloc()分配的指針進行設備內(nèi)存讀寫操作；cudaMalloc()分配的指針傳遞給在設備上執(zhí)行的函數(shù)；但是不可以在主機代碼中使用cudaMalloc()分配的指針進行主機內(nèi)存讀寫操作（即不能進行解引用）。cudaFree與c語言中的free()函數(shù)一樣，只是此函數(shù)釋放的是cudaMalloc()分配的內(nèi)存。

cudaMemcpy與c語言中的memcpy函數(shù)一樣，只是此函數(shù)可以在主機內(nèi)存和GPU內(nèi)存之間互相拷貝數(shù)據(jù)。此外與C中的memcpy()一樣，以同步方式執(zhí)行，即當函數(shù)返回時，復制操作就已經(jīng)完成了，并且在輸出緩沖區(qū)中包含了復制進去的內(nèi)容。相應的有個異步方式執(zhí)行的函數(shù)cudaMemcpyAsync().

該函數(shù)第一個參數(shù)是目的指針，第二個參數(shù)是源指針，第三個參數(shù)是復制內(nèi)存的大小，第四個參數(shù)告訴運行時源指針，這是一個什么類型的指針，即把內(nèi)存從哪里復制到哪里。第四個參數(shù)可以選用以下的形式：參考

cudaMemcpyHostToDevice：從主機復制到設備；cudaMemcpyDeviceToHost：從設備復制到主機；cudaMemcpyDeviceToDevice：從設備復制到設備；cudaMemcpyHostToHost：從主機復制到主機；

調(diào)用的核函數(shù)前面已經(jīng)介紹過，此處不再贅述。還有其他函數(shù)大家可以自行參考官網(wǎng)的API介紹。

大家也可以看《GPU高性能編程CUDA實戰(zhàn)》這本書

opencv中cuda庫安裝與使用

除了直接使用cuda，opencv也有相應的cuda的庫，調(diào)用cuda從而在opencv實現(xiàn)gpu加速功能。這就需要安裝CUDA，并從OpenCV編譯安裝時候安裝好對應cv::cuda庫。

首先要使用cv::cuda里面對應的cv::cuda::add 、cv::cuda::multiply，我們需要安裝opencv要配合OpenCV Contrib庫。

OpenCV Contrib庫是非官方的第三方開發(fā)擴充庫。通過這個庫，我們能使用如dnn、相機標注、3D成像、ArUco、物體追蹤，甚至是需付費的SURF、SIFT特征點提取算法

安裝OpenCV Contrib也是需要和opencv進行版本匹配的。大家可以在下面位置進行對應版本下載：

contrib庫：https://github.com/opencv/opencv_contrib/tagsopencv版本：https://opencv.org/releases.html

對與我個人而言，opencv版本是 4.4.5

$?pkg-config?opencv4?--modversion?
4.4.5

下載好對應的opencv和opencv_contrib庫，開始進行編譯，和之前單獨opencv編譯類似，解壓下載好的文件，然后在opencv建立一個build目錄，在里面進行cmake 編譯。

在這里給大家在注解一些相關選項的含義：

?cmake?-D?CMAKE_BUILD_TYPE=Release
????-D?ENABLE_CXX11=ON
??????????-D?CMAKE_INSTALL_PREFIX=/usr/local
??????????-D?WITH_CUDA=ON
???????????-D?CUDA_ARCH_BIN=${cuda_compute}
???????????-D?CUDA_ARCH_PTX=""
???????????-D?ENABLE_FAST_MATH=ON
???????????-D?CUDA_FAST_MATH=ON
???????????-D?WITH_CUBLAS=ON
????????????-D?WITH_LIBV4L=ON
????????????-D?WITH_GSTREAMER=ON
?????????????-D?WITH_GSTREAMER_0_10=OFF
?????????????-D?WITH_QT=ON
?????????????-D?WITH_OPENGL=ON
?????????????-D?CUDA_NVCC_FLAGS="--expt-relaxed-constexpr"?
?????????????-D?CUDA_TOOLKIT_ROOT_DIR=/usr/local/cuda-11.6
?????????????-D?WITH_TBB=ON
?????????????-D?OPENCV_EXTRA_MODULES_PATH=/home/lyn/Documents/application/opencv_contrib-4.5.5/modules??????????../

CUDA_ARCH_BIN=${cuda_compute} 顯卡算力，大家可以對照我最后給算力查詢自行查詢，也可以進行系統(tǒng)自己查找匹配版本。

WITH_QT=ON 這是是QT的編譯選項 大家可以按照自己需要選擇是否用QT進行編譯開與關。

CUDA_TOOLKIT_ROOT_DIR 這是自己電腦安裝的cuda 版本

OPENCV_EXTRA_MODULES_PATH=/home/lyn/Documents/application/opencv_contrib-4.5.5/modules 這是解壓好后的opencv_contrib目錄

安裝時候可能會遇到如下問題：

In?file?included?from?/home/lyn/Documents/application/opencv-4.5.5/build/modules/python_bindings_generator/pyopencv_custom_headers.h:7,
?????????????????from?/home/lyn/Documents/application/opencv-4.5.5/modules/python/src2/cv2.cpp:88:
/home/lyn/Documents/application/opencv_contrib-4.5.5/modules/phase_unwrapping/misc/python/pyopencv_phase_unwrapping.hpp:2:13:?error:?‘phase_unwrapping’?in?namespace?‘cv’?does?not?name?a?type
????2?|?typedef?cv::phase_unwrapping::HistogramPhaseUnwrapping::Params?HistogramPhaseUnwrapping_Params;

打開-D ENABLE_CXX11=ON 打開C++11的選項 （這就是上面命令使用C++11的原因）

最后編譯成功：

sudo make install這個時候opencv中的cuda庫就安裝好了

介紹一個簡單例子：

CMakeLists.txt文件

#?聲明要求的?cmake?最低版本
cmake_minimum_required(?VERSION?2.8?)
#?聲明一個?cmake?工程
project(opencv_cuda)
#?設置編譯模式
set(?CMAKE_BUILD_TYPE?"Debug"?)
set(?CMAKE_CXX_FLAGS?"-std=c++11")
#添加OPENCV庫
find_package(OpenCV?REQUIRED)
#添加OpenCV頭文件
include_directories(${OpenCV_INCLUDE_DIRS})
#顯示OpenCV_INCLUDE_DIRS的值
message(${OpenCV_INCLUDE_DIRS})
add_executable(target?test_cuda_cv.cpp)
#?將庫文件鏈接到可執(zhí)行程序上
target_link_libraries(target??${OpenCV_LIBS}?)?

test_cuda_cv.cpp文件

#include?<iostream>
#include?"opencv2/opencv.hpp"
#include?<opencv2/core/cuda.hpp>?
#include?<opencv2/cudaarithm.hpp>
#include?<opencv2/core/version.hpp>

int?main?(int?argc,?char*?argv[])
{
????//Read?Two?Images?
????cv::Mat?h_img1?=?cv::imread(?"/home/lyn/Documents/work-data/test_code/opencv/learn_code/"
??????"opencv_tutorial_data-master/images/sp_noise.png");
????cv::Mat?h_img2?=?cv::imread(?"/home/lyn/Documents/work-data/test_code/opencv/learn_code/"
??????"opencv_tutorial_data-master/images/sp_noise.png");??????
????//?cv::Mat?h_img2?=?cv::imread("/home/lyn/Documents/work-data/test_code/opencv/learn_code/c++/build/filter_line.png");
??
??//?cv::namedWindow(?"1",?cv::WINDOW_FREERATIO);?
??//?cv::imshow("1",h_img1);??

??//?cv::namedWindow(?"2",?cv::WINDOW_FREERATIO);?
??//?cv::imshow("2",h_img2);??

????cv::Mat?h_result1;


????#if?0
????//?cv::add(h_img1,?h_img2,h_result1);//加法操作?api
????cv::multiply(h_img1,cv::Scalar(2,?2,?2),h_result1);//乘法操作?api

????#else
????//?定義GPU數(shù)據(jù)
????cv::cuda::GpuMat?d_result1,d_img1,?d_img2;
????
????//?CPU?到?GPU???
????d_img1.upload(h_img1);
????d_img2.upload(h_img2);
????
????//?調(diào)用GPU接執(zhí)行?mat?substract
????//?cv::cuda::add(d_img1,?d_img2,d_result1);
????cv::cuda::multiply(d_img1,cv::Scalar(2,?2,?2),d_result1);//乘法操作?api

????//?gpu?結(jié)果?到?cpu
????d_result1.download(h_result1);

????#endif?
?

????//?顯示，保存
??//?cv::namedWindow(?"Image1",?cv::WINDOW_FREERATIO);?
????cv::imshow("Image1?",?h_img1);

??//?cv::namedWindow(?"Image2",?cv::WINDOW_FREERATIO);?
????cv::imshow("Image2?",?h_img2);

??//?cv::namedWindow(?"Result_Subtraction",?cv::WINDOW_FREERATIO);?
????cv::imshow("Result_Subtraction?",?h_result1);
????cv::imwrite("/home/lyn/Documents/work-data/test_code/opencv/learn_code/c++/result_add.png",?h_result1);
????cv::waitKey();
????return?0;
}

大家也可以使用opencv中的其他cuda庫，具體函數(shù)使用在下面

官方鏈接：https://docs.opencv.org/4.5.5/d1/d1e/group__cuda.html

大家可以在當前頁面的版本選擇中，選擇和自己一致的opencv版本，進行參考：

例如我查詢cv::cuda::add函數(shù)，就在 Operations on Matrices -> Per-element Operations

也可以看具體的數(shù)據(jù)結(jié)構(gòu)含義：例如 cv::cuda::GpuMat

官方的解釋是：具有引用計數(shù)的 GPU 內(nèi)存的基本存儲類。GpuMat類似于Mat，不過管理的是GPU內(nèi)存，也就是顯存。GpuMat和Mat相比，存在以下限制：

• 不支持任意尺寸（只支持2D，也就是二維數(shù)據(jù)）沒有返回對其數(shù)據(jù)的引用的函數(shù)（因為 GPU 上的引用對 CPU 無效）不支持c++的模板技術。

總之是很方便的，大家可以按照自己需要進行查詢信息。

附錄：這是官網(wǎng)提供算力表，大家可以參考選擇自己的要求算力的顯卡。

https://developer.nvidia.com/zh-cn/cuda-gpus

結(jié)語

這就是我自己的一些cuda的使用分享，因為篇幅所限，下一篇，我們聊聊cuda實際使用gpu加速和cpu使用的對比，以及它兩使用場景分析。如果大家有更好的想法和需求，也歡迎大家加我好友交流分享哈。

作者：良知猶存，白天努力工作，晚上原創(chuàng)公號號主。公眾號內(nèi)容除了技術還有些人生感悟，一個認真輸出內(nèi)容的職場老司機，也是一個技術之外豐富生活的人，攝影、音樂 and 籃球。關注我，與我一起同行。