Cuda 9.0 範例程式 使用 for 迴圈複製陣列

現在9版的安裝 cuda 好像沒有什麼難度了，順對弄對先安裝好 VisualStudio 然後再安裝 cuda 就一路下一步一直按到底就好了。

然後開啟VC就有專案可以開啟了

範例程序有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;
}

​