CUDA 内核可以是虚函数吗？

Question

这个问题很简单，但让我概述一下我的框架。我有一个抽象 class AbstractScheme 表示一种计算类型（一种方程的离散化，但这并不重要）。每个实现都必须提供一个 return 方案名称的方法，并且必须实现一个受保护的函数，即 CUDA 内核。基本抽象 class 提供了一个 public 调用 CUDA 内核的方法和 return 内核完成所花费的时间。

class AbstractScheme
{
public:
    /**
     * @return The name of the scheme is returned
     */
    virtual std::string name() const =0;

    /**
     * Copies the input to the device,
     * computes the number of blocks and threads,
     * launches the kernel,
     * copies the output to the host,
     * and measures the time to do all of this.
     *
     * @return The number of milliseconds to perform the whole operation
     *         is returned
     */
    double doComputation(const float* input, float* output, int nElements)
    {
        // Does a lot of things and calls this->kernel().
    }

protected:
    /**
     * CUDA kernel which does the computation.
     * Must be implemented.
     */
    virtual __global__ void kernel(const float*, float*, int) =0;
};

我也有几个这个基础的实现 class。但是当我尝试使用 nvcc 7.0 进行编译时，我收到此错误消息，指的是我在 AbstractScheme 中定义函数 kernel 的行（上面清单中的最后一行）：

myfile.cu(60): error: illegal combination of memory qualifiers

我找不到任何资源说内核不能是虚函数，但我觉得这就是问题所在。你能解释一下这背后的理由吗？我清楚地理解 __device__ 函数不能是虚函数的方式和原因（虚函数是指向存储在对象中的实际 [host] 函数的指针，你不能从设备代码中调用这样的函数），但我不确定 __global__ 函数。

编辑：问题中我划掉的部分是错误的。请查看评论以了解原因。

Answer 1

内核不能是 CUDA 对象模型中类的成员，无论是否是虚拟的。这就是编译错误的原因，即使从编译器发出的错误消息来看不是特别明显。

CUDA 内核可以是虚函数吗？

Can CUDA kernels be virtual functions?

c++

polymorphism

cuda

virtual-functions