[CUDA] CudaFilter

使用 cudaFilter 的範例 (須 include cudafilters.hpp)

[CUDA+OpenCV] GpuMat 結合 kernel 運算

OpenCV 的 GpuMat 類別, 可以在 device 的 global memory 建立 Mat 物件, 可以被 kernel 函數直接存取, 以下以 sobel 運算子為例, 將 srcMat (位於host memory) 載入影像後, 轉換至 srcGMat (位於device memory), 經處理後放至 destGMat (位於device memory), 然後下載回 destMat (位於host memory)再顯示出來.
程式中使用 OpenCV 定義於 cuda_types.hpp 的 PtrStep 的樣板結構, 代入定義於 WIndows.h 的 byte, 這樣比較方便. 如果要用 cuda 編譯器 nvcc 內建的  uchar1 , 查看 vector_types.h, 它是一個結構 (向量):

[CUDA] 與 OpenCV 共舞

取得 CUDA 版的 OpenCV 後, 使用 OpenCV 前, 先建立環境變數 OPENCV_DIR, 使其指向 opencv 目錄 ( C:\opencv\build\ ):

[CUDA] OpenCV CUDA 程式庫

OpenCV  現在有支援 CUDA 與 OpenCL, 但是官網預先編譯過的程式庫並未包含 CUDA, 所以要自行從原始檔進行編譯.
1. 先從 OpenCV官網下載原始檔, 然後解壓縮.
2. 下載 CMake 安裝檔, 並完成安裝.

[CUDA] 更新Visual Studio後無法編譯的問題

由於 CUDA 對 Visual Stidio 的 Integration 會檢查版本, 如果安裝 CUDA 之後更新升級 Visual Stdio, 可能導致因版本錯誤而出現以下訊息:

#error -- unsupported Microsoft Visual Studio version! Only the versions 2012, 2013, 2015 and 2017 are supported!

[CUDA] 平行化之 Scan 演算法

影像處理 histogram 等化處理時, 把 histogram 視為機率分布函數 (pdf), 經常將其積分為累積分布函數 cdf, 例如:

        pdf = { 1, 2, 3, 4, 5, 6, 7, 8}

        cdf = { 1, 3, 6, 10, 15, 21 28, 36}

程式碼為:

[CUDA] 平行化之 Reduce 演算法

考慮一個典型的總和運算 :

int sum = 0;      for (i = 1; i < N; i++) {          sum += a[i];      }

[CUDA] 多重讀取與寫入的互斥問題

如果 kernel 執行時, 不同的 thread 對同一個變數進行讀取與寫入, 會發生甚麼事呢? 當不同thread被同時執行時, 同時去讀取同一個變數, 得到相同的值, 如果把此值運算後同時又寫回去, 最後的值就不一定是多少了, 以下程式開啟 BlockSize*ThreadSize 個 thread 去執行把 global memory 陣列變數加1的動作,程式碼如下 :

[CUDA] 測量效能

這裡使用兩種方式:

1. 使用 windows的 PerformanceCounter, 只要 include <Windows.h> 即可使用
2. 使用 cudaEvent

arraySize受限於host與gpu的記憶體容量, 可自行調整測試, repeat為重複執行次數. 如果在 main() 中重複呼叫 addWithCuda(), 會因為記憶體搬移花費大量時間, 反而可能比CPU還慢, 因此直接在 addWithCuda() 中重複呼叫 kernel去計算. 資料陣列可宣告於 globalㄝ如果放在local 最好宣告 static, 否則  arraySize 太大會造成 stack overflow.

[CUDA] 檢視 CUDA 性能

CudaInfo(): 列印Cuda相關能力

[CUDA] host memory與device memory

以下程式碼展示 host memory與device memory之間的關係, 變數命名 host 開頭表示是在 host memory, dev 開頭表示是在 device memory. 一開始初始化 host_a[], 然後用cudaMemcpy() 複製至 dev_a[], 平行呼叫 kernel 函數 shared() 後, 將 dev_a[] 複製至 x[], x是shared memory, 可在同一個 block 內的所有 thread 使用.

MR與實際地震的關係

RMT的MR值只看相關性, 與震動的實際絕對大小無關, 小震動也有可能造成高MR值. 照原本RMT的設計理念, 地震發生MR應該會升高, 如果把地震發生看成一個變數, MR高低看成一個變數, 組合起來有4種狀況:
1. 有地震, MR高 : 這個RMT已經解決了
2. 有地震, MR低 : 目前觀察, 小區域小規模地震時, 震波沒有傳到另一個測站, 容易產生這個現象, 經驗上大概規模4以上就比較少這個問題, 規模3.5偶而還是會看到
3. 沒地震, MR低 : 這應該就是正常的雜訊
4. 沒地震, MR高 : 可能是真的有小震, 但因為震幅過小但是有相關性, 也可能是隨機雜訊剛好有相關性, 這可以用時間排除, 真是隨機訊號的話, 相關性只會出現在瞬間, 不會持續達幾十分鐘之久, 這樣並不合理

[CUDA] 平行處理基礎 : Block 與 Thread

在 CUDA 中, 以軟體觀點來看, 一個 grid 由許多 block 組成, 一個 block 由許多 thread 組成, 每一個 kernel 函數執行時, 可以從 threadIdx 與 blockIdx 取得編號, 但是並沒有 gridIdx, 這表示每一個平行處理的執行, 是在同一個 grid 下進行. BlockSize 與 ThreadSize 的數量受限於 GPU 的硬體架構, 平行處理的函數呼叫語法是:

在 C# 呼叫 CUDA 的方法

最近開始研究CUDA，準備做演算法加速， 因此生出這篇筆記。因  CUDA 是叫用 nvcc 進行編譯，CUDA 函數無法被C# 這種 managed code 直接呼叫，在同一個 CUDA 專案中只能用C或C++來呼叫 CUDA，因此必須在 CUDA 專案中加一個 C/C++ 的 wrapper 函數，將 CUDA 專案包裝成 C 語言的DLL檔，然後在 C# 中用 DllImport 呼叫 C 函式來轉給 CUDA 計算。以下實作以 Visual Studio 2017 Community CUDA 9.2 SDK為例，完整程式碼在https://github.com/ghostyguo/CudaDotNet。