CPU到GPU
在CUDA程序中,将数据从CPU传输到GPU,或者从GPU传输到CPU的时候,需要调用底层的内存拷贝函数。当有很多不同类型的数据的时候,这个过程会非常繁琐。于是,我专门写了个内存拷贝的模板函数,使这个过程变的十分方便。
template <typename T> T* valueHostToDevice(T *value, const int &num = 1, bool isDelete = false){
T *devValue;
cudaMalloc((void**)&devValue, num * sizeof(T));
cudaMemcpy(devValue, value, num * sizeof(T), cudaMemcpyHostToDevice);
if(isDelete){
if(num == 1){delete value;}
else{delete[] value;}
}
return devValue;
}
以上的函数能够将数据从CPU拷贝到GPU,而且能够支持任意类型的数据,比如整型int、浮点型float或自定义的类型,同时既能拷贝单个对象,也能拷贝一个对象数组,最后还可以销毁CPU内存数据,其调用方法也是非常方便
// 任意类型的数组指针
A *m_a = new A[3];
// 从CPU的m_a拷贝到GPU的a
A *a = valueHostToDevice(m_a, 3);
以上是带有返回值的函数,也可以写成是不带返回值的函数
template <typename T> void valueHostToDevice(T &value, T **devValue){
T *hostValue;
cudaHostAlloc((void**)&hostValue, sizeof(T), cudaHostAllocDefault);
cudaMalloc((void**)devValue, sizeof(T));
*hostValue = value;
cudaMemcpy(*devValue, hostValue, sizeof(T), cudaMemcpyHostToDevice);
cudaFreeHost(hostValue);
}
// 调用
valueHostToDevice(m_a, &a);
以上这个函数就有点繁琐了。为了GPU的开辟内存空间,需要使用cudaMalloc函数操作GPU指针a。cudaalloc函数需要传入指针a的地址,而直接传入a的话,这是作为形参的,所以函数里的指针是个临时变量,并不是原始的指针a的地址,所以在函数中直接传入指针a是无法操作的。所以需要定义二级指针,再让这个二级指针指向a指针,才能对其进行操作,所调用时需要传入的形参是&a。
GPU到CPU
类似的还有从GPU传到CPU的过程,其调用方法也是类似的。
template <typename T> T* valueDeviceToHost(T *devValue, const int &num = 1){
T *hostValue;
cudaHostAlloc((void**)&hostValue, num * sizeof(T), cudaHostAllocDefault);
cudaMemcpy(hostValue, devValue, num * sizeof(T), cudaMemcpyDeviceToHost);
return hostValue;
}
template <typename T> void valueDeviceToHost(T *devValue, T &value){
T *hostValue;
cudaHostAlloc((void**)&hostValue, 1 * sizeof(T), cudaHostAllocDefault);
cudaMemcpy(hostValue, devValue, 1 * sizeof(T), cudaMemcpyDeviceToHost);
value = *hostValue;
cudaFreeHost(hostValue);
}
二维数组
以上只是拷贝一维数组,同时还可以拷贝二维数组。
template <typename T> T** array2DHostToDevice(T *hostData, T **devData,
const int &rowNum, const int &colNum){
// 开辟内存
T **hostArray;
cudaHostAlloc((void**)&hostArray, rowNum * sizeof(T*), cudaHostAllocDefault);
T **devArray;
cudaMalloc((void**)&devArray, rowNum * sizeof(T*));
size_t pitch;
cudaMallocPitch(devData, &pitch, colNum * sizeof(T), rowNum);
cudaMemset2D(*devData, pitch, 0, colNum * sizeof(T), rowNum);
// 设置二维数组的指针
for(int i = 0; i < rowNum; i++){
hostArray[i] = (T*)((float*)(*devData) + i*pitch / sizeof(float));
}
// 复制数据
if(hostData != NULL)
cudaMemcpy2D(*devData, pitch, hostData, colNum * sizeof(T), colNum * sizeof(T), rowNum, cudaMemcpyHostToDevice);
cudaMemcpy(devArray, hostArray, rowNum * sizeof(T*), cudaMemcpyHostToDevice);
cudaFreeHost(hostArray);
return devArray;
}
// 调用
vec **image2D = NULL, *gpu_image2DData = NULL;
vec *cpu_image2DData = NULL;
image2D = array2DHostToDevice(cpu_image2DData, &gpu_normalData, height, width);
template <typename T> void array2DHostToDevice(T ***devArray,
T *hostData, T **devData,
const int &rowNum, const int &colNum){
// 开辟内存
T **hostArray;
cudaHostAlloc((void**)&hostArray, rowNum * sizeof(T*), cudaHostAllocDefault);
cudaMalloc((void**)devArray, rowNum * sizeof(T*));
size_t pitch;
cudaMallocPitch(devData, &pitch, colNum * sizeof(T), rowNum);
cudaMemset2D(*devData, pitch, 0, colNum * sizeof(T), rowNum);
// 设置二维数组的指针
for(int i = 0; i < rowNum; i++){
hostArray[i] = (T*)((float*)(*devData) + i*pitch / sizeof(float));
}
// 复制数据
cudaMemcpy2D(*devData, pitch, hostData, colNum * sizeof(T), colNum * sizeof(T), rowNum, cudaMemcpyHostToDevice);
cudaMemcpy(*devArray, hostArray, rowNum * sizeof(T*), cudaMemcpyHostToDevice);
cudaFreeHost(hostArray);
}
封装成类之后,可以更加方便。
class MemCopy{
public:
MemCopy(){}
template <typename T> T* hostToDevice(T *value,
const int &num = 1,
bool isDelete = false);
template <typename T> T** hostToDevice(T *hostData,
T **devData,
const int &rowNum,
const int &colNum);
};
template <typename T> T* MemCopy::hostToDevice(T *value,
const int &num,
bool isDelete){
// 任意类型的数组指针
// A *m_a = new A[3];
// 从CPU的m_a拷贝到GPU的a
// A *a = valueHostToDevice(m_a, 3);
T *devValue;
cudaMalloc((void**)&devValue, num * sizeof(T));
cudaMemcpy(devValue, value, num * sizeof(T), cudaMemcpyHostToDevice);
if(isDelete){
if(num == 1){delete value;}
else{delete[] value;}
}
return devValue;
}
template <typename T> T** MemCopy::hostToDevice(T *hostData,
T **devData,
const int &rowNum,
const int &colNum){
// 调用
// vec **image2D = NULL, *gpu_image2DData = NULL;
// vec *cpu_image2DData = NULL;
// image2D = array2DHostToDevice(cpu_image2DData, &gpu_normalData, height, width);
// 开辟内存
T **hostArray;
cudaHostAlloc((void**)&hostArray, rowNum * sizeof(T*), cudaHostAllocDefault);
T **devArray;
cudaMalloc((void**)&devArray, rowNum * sizeof(T*));
size_t pitch;
cudaMallocPitch(devData, &pitch, colNum * sizeof(T), rowNum);
cudaMemset2D(*devData, pitch, 0, colNum * sizeof(T), rowNum);
// 设置二维数组的指针
for(int i = 0; i < rowNum; i++){
hostArray[i] = (T*)((float*)(*devData) + i*pitch / sizeof(float));
}
// 复制数据
if(hostData != NULL)
cudaMemcpy2D(*devData, pitch, hostData, colNum * sizeof(T), colNum * sizeof(T), rowNum, cudaMemcpyHostToDevice);
cudaMemcpy(devArray, hostArray, rowNum * sizeof(T*), cudaMemcpyHostToDevice);
cudaFreeHost(hostArray);
return devArray;
}
int main(void){
Timing t = Timing();
int *c_a = new int[100000000];
for(int i = 0; i < 10000; i++){
for(int j = 0; j < 10000; j++){
c_a[i*10000+j] = 10;
}
}
MemCopy c = MemCopy();
t.tic();
// Do Something
int *g_a;
int **a = c.hostToDevice(c_a, &g_a, 10000, 10000);
t.toc();
delete[] c_a;
cudaFree(g_a);
cudaFree(a);
return 0;
}