C++使用类调用CUDA核函数_C

概述正如CUDA C所称，CUDA对C语言进行了很好的扩展，直接使用C语言可以非常简单方便的调用CUDA核函数。但是当想使用C++的类成员函数直接调用核函数是不可行的，第一，核函数不能作为类的成员函数，第二，C++的cpp文件和CUDA的cu文件分别经由g++和nvcc编译，当两种代码混合就会编译出错。因而C++的类和CUDA结合使用需要进行一层封装，借用两个数组相加的例子说明，主要过程如下：

正如CUDA C所称，CUDA对C语言进行了很好的扩展，直接使用C语言可以非常简单方便的调用CUDA核函数。但是当想使用C++的类成员函数直接调用核函数是不可行的，第一，核函数不能作为类的成员函数，第二，C++的cpp文件和CUDA的cu文件分别经由g++和nvcc编译，当两种代码混合就会编译出错。

因而C++的类和CUDA结合使用需要进行一层封装，借用两个数组相加的例子说明，主要过程如下：

本项目包括4文件，如上图所示，分别为：add.h、add.cpp、kernel.cuh、kernel.cu。在add.h中封装一个函数AddNum()，调用kernel.cuh中的函数

AddKernel(int *a,int *b,int *c,int DX)，然后在kernel.cu文件中使用AddKernel(...)调用相加核函数Add(int *a,int@H_301_30@ DX)

1、add.h文件定义了一个CTest的类，包括3个指针（数组）、4个函数。

#pragma once@H_301_30@#include "kernel.cuh"@H_301_30@#include <iostream>using namespace@H_301_30@ std;#define DX 200class@H_301_30@ CTest{public@H_301_30@:    int *@H_301_30@a;    int *@H_301_30@b;    int *@H_301_30@c;    voID@H_301_30@ SetParameter();    voID@H_301_30@ AddNum();    voID@H_301_30@ Show();    voID@H_301_30@ Evolution();};voID@H_301_30@ CTest::SetParameter(){    cudamallocManaged(&a,sizeof(int) *@H_301_30@ DX);    cudamallocManaged(&b,sizeof(int) *@H_301_30@ DX);    cudamallocManaged(&c,sizeof(int) *@H_301_30@ DX);    for (int f = 0; f<DX; f++@H_301_30@)    {        a[f] =@H_301_30@ f;        b[f] = f + 1@H_301_30@;    }}voID@H_301_30@ CTest::AddNum(){    AddKernel(a,b,c,DX);}voID@H_301_30@ CTest::Show(){    cout << " a     b    c"  <<@H_301_30@ endl;    for (int f = 0; f<DX; f++@H_301_30@)    {        cout << a[f] << " + " << b[f] << "  = " << c[f] <<@H_301_30@ endl;    }}voID@H_301_30@ CTest::Evolution(){    SetParameter();    AddNum();    Show();}

2、add.cpp文件执行主函数，创建一个CTest的对象cTest，然后调用Evolution执行相加 *** 作。

#include "add.h"voID@H_301_30@ main(){    CTest cTest;    cTest.Evolution();    system("pause"@H_301_30@);}

3、kernel.cuh文件定义一个接口函数AddKernel(int *a,int DX);

#include "cuda_runtime.h"@H_301_30@#include "device_launch_parameters.h"voID AddKernel(int *a,int *b,int *c,int DX);

4、kernel.cu包括相加Add核函数以及调用核函数的封装函数AddKernel.

#include "kernel.cuh"@H_301_30@__global__ voID Add(int *a,int@H_301_30@ DX){    int f = blockIDx.x*blockDim.x +@H_301_30@ threadIDx.x;    if (f >= DX) return@H_301_30@;    c[f] = a[f] +@H_301_30@ b[f];}voID AddKernel(int *a,int@H_301_30@ DX){    dim3 dimBlock = (128@H_301_30@);    dim3 dimgrID = ((DX + 128 - 1) / 128@H_301_30@);    Add << <dimgrID,dimBlock >> >@H_301_30@ (a,DX);    cudaDeviceSynchronize();}