Cuda 9.0 範例程式 使用 for 迴圈複製陣列
現在9版的安裝 cuda 好像沒有什麼難度了,順對弄對先安裝好 VisualStudio 然後再安裝 cuda 就一路下一步一直按到底就好了。
範例程序有for迴圈可以參考,這裡我把它簡化了一下一些流程並加上時間監控
首先先是計時函式庫
/*****************************************************************
Name : Timer.hpp
Date : 2017/12/19
By : CharlotteHonG
Final: 2017/12/26
*****************************************************************/
#pragma once
#include <iostream>
#include <string>
#include <ctime>
class Timer {
public:
Timer(std::string name=""): name(name){
startTime = clock();
}
operator double() {
return time;
}
public:
void start() {
startTime = clock();
flag=0;
}
void end() {
finalTime = clock();
time = (double)(finalTime - startTime)/CLOCKS_PER_SEC;
}
void print(std::string name="") {
if (flag==0) {
flag = 1;
end();
}
if(name=="") {
name=this->name;
}
if(priSta) {
std::cout << "#" << name << ", " << " time = " << time << "s" << std::endl;
}
}
private:
std::string name;
clock_t startTime;
clock_t finalTime;
double time;
bool flag = 0;
public:
bool priSta = 1;
};
再來是控管 cuda 記憶體的類別,他可以用來自動要求空間並複製進去以及在程序結束時自動解構,不需要再 free() cuda記憶體。
由於這個有寫樣板所以沒法簡單把實作拆開,建議就直接寫在cuh裡面省事吧~
/*****************************************************************
Name : cudaData.cuh
Date : 2017/12/19
By : CharlotteHonG
Final: 2017/12/19
*****************************************************************/
// Cuda 記憶體自動管理程序
template <class T>
class CudaData {
public:
CudaData(){}
CudaData(size_t size){
malloc(size);
}
CudaData(T* dataIn ,size_t size): len(size){
memcpyInAuto(dataIn, size);
}
~CudaData(){
if(gpuData!=nullptr) {
cudaFree(gpuData);
gpuData = nullptr;
len = 0;
}
}
public:
void malloc(size_t size) {
this->~CudaData();
len = size;
cudaMalloc((void**)&gpuData, size*sizeof(T));
}
void memcpyIn(T* dataIn ,size_t size) {
if(size > len) {throw out_of_range("memcpyIn input size > curr size.");}
cudaMemcpy(gpuData, dataIn, size*sizeof(T), cudaMemcpyHostToDevice);
}
void memcpyInAuto(T* dataIn ,size_t size) {
malloc(size);
memcpyIn(dataIn, size);
}
void memcpyOut(T* dataIn ,size_t size) {
cudaMemcpy(dataIn, gpuData, size*sizeof(T), cudaMemcpyDeviceToHost);
}
void memset(int value, size_t size) {
if(size>len) {
throw out_of_range("memset input size > curr size.");
}
cudaMemset(gpuData, value, size*sizeof(T));
}
size_t size() {
return this->len;
}
public:
operator T*() {
return gpuData;
}
private:
T* gpuData;
size_t len=0;
};
主程式
再來是主程式了,會用到上面兩個函式庫
#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <iostream>
#include <vector>
using namespace std;
#include "timer.hpp"
#include "CudaData.cuh"
__global__ void cudacopy(float* b, float* a, int size){
// 乾式寫法
const int idx = blockIdx.x*blockDim.x + threadIdx.x;
if(idx<size){
b[idx]=a[idx];
}
// 迴圈寫法
for(int i=threadIdx.x; i<size; i+=blockDim.x){
b[i]=a[i];
}
}
void cpucopoy(float* b, float* a, int size) {
for(int i=0; i<size; ++i){
b[i]=a[i];
}
}
void testCuda(size_t size) {
Timer T;
// 配置主機記憶體
vector<float> img_data(size), cpu_data(size), gpu_data(size);
float* a = img_data.data(); // 原始資料
float* b = cpu_data.data(); // CPU計算後資料
float* c = gpu_data.data(); // GPU輸出回來資料
// 設置初值
float test_val=7;
for(int i=0; i<size; i++){
a[i]=test_val;
}
// 配置顯示記憶體, 載入資料.
T.start();
CudaData<float> gpuDataIn(a, size), gpuDataOut(size);
gpuDataOut.memset(0, size);
T.print(" Cuda Data malloc and copy");
// 網格區塊設定. (與 kernel for 的次數有關)
const size_t blkDim=16;
int grid(size/blkDim+1); // 網格要含蓋所有範圍, 所以除完要加 1.
int block(blkDim); // 區塊設定 16x16.
// Cuda Kernel 執行運算
T.start();
cudacopy<<<grid,block>>>(gpuDataOut, gpuDataIn, size);
T.print(" Cuda-copy");
// 取出GPU資料
gpuDataOut.memcpyOut(c, size);
// CPU 執行運算
T.start();
cpucopoy(b, a, size);
T.print(" Cpu-copy");
// 測試
bool f=0;
for(size_t i = 0; i < size; i++) {
if(c[i] != b[i]) {f=1;}
}
// 測試報告
if(f==0) {
cout << "test ok" << endl;
} else {
cout << "test Error" << endl;
}
}
int main(){
Timer T;
testCuda(1000);
T.print("ALL time.");
return 0;
}