2011-06-14 99 views
7

时,我一直在试图创建模板的内核,但我已经遇到了一些麻烦称他们在我的计划。我有一个Matrix<T>模板类,和里面定义的一些方法问题调用模板CUDA内核

Matrix.h:

template <typename T> class Matrix { 
    ... 
    void sum(Matrix<T>& m1, Matrix<T>& m2, Matrix<T>& sum); 
    ... 
} 

#include "Matrix.cu" 

Matrix.cu:

#include "MatrixKernel.h" 

template<typename T> void Matrix<T>::sum(const Matrix<T>& m, Matrix<T>& sum) { 
    ... 
    sumKernel<T><<<dimGrid, dimBlock>>>(Matrix<T> m1, Matrix<T> m2, Matrix<T> sum) 
    ... 
} 

MatrixKernel.h:

template<typename T> __global__ void sumKernel(const Matrix<T> m1, const Matrix<T> m2, Matrix<T> sum) { 
... 
} 

问题是,当我从sum中调用sumKernel时,编译器给了我下面的错误:

error C2059: syntax error : '<' 

有人知道发生了什么吗?在包含sumKernel调用之前,代码就可以正常编译了。

谢谢。

+0

我不知道你可以使用CUDA和C++(!)。琐碎的建议:尝试在''和'<<<'之间放置一个空格,以防将它们放在一起导致解析问题。 – Rup 2011-06-14 10:43:33

+0

编译器是否告诉你哪一行出错?在cuda模板代码中有很多 2011-06-14 10:45:34

+0

@Rup:是的,你可以。您甚至可以将对象作为参数传递给内核(只要您将感兴趣的数据复制到设备内存)。我也会尝试你的建议。 @Bomadeno:错误在执行内核调用的线上。 – Renan 2011-06-14 15:42:24

回答

6

所以,看来你确实有一个奇怪的#include,导致代码过得去错误编译器编译。通过为cuda头文件使用.cu.h来区分gpu头文件和cpu头文件。确保只有 NVCC编译.cu.cu.h文件。 Cuda文件不应该包含在cpp文件中。内核和内核调用应该在.cu.cu.h文件,这些文件不应该在任何地方CPPS包括。

由于您的.cu正在包含在由主机编译器编译的头文件中,主编译器最终会碰到令牌<<< - 它无法识别。它可能确实理解了令牌<<,因此它会消耗此功能,从而产生意外的<

这里做的事情应该工作的另一种方式(没有尝试过,但它类似于我们使用的代码)

(注意,这可能会奏效,但它也可能不是解决问题的正确方法。我的老板不喜欢它作为解决方案,并且倾向于为每个变体添加实现)

底层问题似乎是主机和设备代码之间缺乏区别。我要离开的细节,并且在我的解决方案 - 比如复制结果,并从设备,和实施等

我试图解决的是,给定一个结构的问题,你怎么能模板它在主机和设备上都使用?

我会在这两个类型和实现细节模板Matrix.h

template <typename T, typename Implementation<T> > class Matrix { 
    void sum(Matrix<T>& m1, Matrix<T>& m2, Matrix<T>& sum) 
    { 
     Implementation.sumImp(m1, m2, sum); 
    } 
} 

主机实现,HostMatrixSum.h会做的事情上的CPU:

#include "Matrix.h" 

template <typename T> struct HostMatrixSum 
{ 
    void sumImp(Matrix<T>& m1, Matrix<T>& m2, Matrix<T>& sum) 
    { 
     ... 
    } 
} 

虽然GpuMatrixSum.cu.h将上传矩阵,做之和恢复的结果:

#include "Matrix.h" 

template <typename T> struct GpuMatrixSum 
{ 
    template<typename T> __global__ void sumKernel(const Matrix<T> m1, const Matrix<T> m2, Matrix<T> sum) 
    { 
     ... 
    } 

    void sumImp(Matrix<T>& m1, Matrix<T>& m2, Matrix<T>& sum) 
    { 
     ... 
     sumKernel<T> <<< dimGrid, dimBlock >>> (m1,m2); 
     ... 
    } 
} 

然后当我们从主机代码中使用Matrix时,我们在主机总和实现上进行模板化,并且从不需要查看任何cuda细节:

#include "Matrix.h" 
#include "HostMatrixSum.h" 

Matrix<int, HostMatrixSum> m1 = Matrix<int>(...); 
Matrix<int, HostMatrixSum> m2 = Matrix<int>(...); 
Matrix<int, HostMatrixSum> result; 
Matrix.sum(m1,m2,result); 

如果我们在GPU上的合作,我们可以使用GPU加速实施总和:

#include "Matrix.h" 
#include "GpuMatrixSum.cu.h" 

Matrix<int, GpuMatrixSum> m1 = Matrix<int>(...); 
Matrix<int, GpuMatrixSum> m2 = Matrix<int>(...); 
Matrix<int, GpuMatrixSum> result; 
Matrix.sum(m1,m2,result); 

希望对你有用!

+0

我要试一试。但同时在MatrixKernel.h中,例如编译器不会抱怨__global__关键字(它只能表示NVCC正在编译它,对吧?) 另一件事:如果你说实话是问题,我会在哪里实施总和方法?如果我没有在“Matrix.h”中写入'#include'Matrix.cu“',将会出现链接错误,因为模板必须在同一个文件中声明和定义... – Renan 2011-06-14 15:53:09

+0

我认为你是对的,我不得不忘记将Matrix作为模板类来实现,因为我无法以这种方式实现它。如果我在头文件中包含一个.cu文件,那么包含Matrix头文件的其他所有文件都将包含.cu,甚至是.cpp文件,这将不可避免地导致编译错误。使用模板内核是可以的,但是由于我刚刚解释的内容,使得称为模板的C++方法也不可行。毕竟它有点缠绕...... – Renan 2011-06-15 06:53:49

+0

因为MatrixKernel.h只包含在一个cu文件中,所以只有nvcc才包含它。如果您将MatrixKernel.h包含在主机cpp文件中,我怀疑它会崩溃。我将cuda命名为特定头文件.cu.h,以澄清它仅适用于.cu文件。我正在回答'如何去做'问题的一部分 - 试图找出一个优雅的解决方案:) – 2011-06-15 10:16:40