2014-01-30 77 views
0

我能夠將普通對象作爲副本傳遞給內核函數。但是,當我將虛擬繼承添加到類層次結構時,我收到一條錯誤消息,指出具有用戶定義的副本構造函數的類不能用作內核啓動的參數。但是,我沒有用戶定義的複製構造函數。所以我想這是因爲虛擬繼承的內部實現實現了一些不同類型的複製構造函數。有人可以解釋虛擬繼承是如何實際實現的嗎?是否有任何解決方法,或者在編寫cuda代碼時絕對沒有辦法使用虛擬繼承?虛擬繼承傳遞給cuda內核函數的對象

的代碼是這樣的:

class Base {...}; 
class ChildA: public virtual Base {...}; 
class ChildB: public virtual Base {...}; 
class GrandChild: public ChildA, public ChildB {...}; 

__global__ void mykernel(Base x) {...} 

int main() { 
    GrandChild x; 
    mykernel<<<1,1>>>(x); 
    return 0; 
} 

編輯: 這是我的猜測:我覺得NVCC只允許默認的拷貝構造函數,因爲在這種情況下,它可以簡單地使用cudaMemcpyAsync的論點推入設備內部的調用堆棧。所以它對編譯器進行了硬編碼,使得只允許使用默認的複製構造函數,但是具有虛擬繼承的對象在內部具有不同類型的複製構造函數,從而觸發了nvcc中的錯誤。不過,我仍然相信nvcc應該有一個簡單的方法允許它,只要nvcc支持虛函數和其他高級C++功能。

+0

在這裏看看它是如何實現的(在gcc中)http://www.phpcompiler.org/articles/virtualinheritance.html;它可能會給你一個第一個想法。只要虛擬功能得到支持,我就不會看到支持虛擬繼承的技術限制的簡單解釋。 – Joky

+0

您正在按值傳遞x。 – IdeaHat

+0

您可能感興趣的[this](http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#virtual-functions) –

回答

1

感謝您的所有意見。根據@RobertCrovella提供的鏈接中的以下兩條語句,我所做的是不允許的。

  1. 不允許作爲參數傳遞給一個__global__函數的類的一個對象與虛擬功能。
  2. 不允許將從虛擬基類派生的類的對象作爲參數傳遞給__global__函數。

與鏈路@Joky提供的,我相信原因是因爲NVCC使用一個簡單的內存拷貝從主機內存參數推到在設備內存中調用堆棧。這就是爲什麼不允許使用非默認的複製構造函數的原因,因爲只有默認的複製構造函數與這種簡單的內存複製行爲一致。

我正在使用的解決方案是打破一個繼承關係並設置一個類型轉換運算符以保留上傳的缺失鏈接,如下所示。這對我很好,因爲所有這些類只是指針的包裝,並且類型轉換足夠高效,即使價格有時必須進行明確的類型轉換,例如在以下調用mykernel2函數的情況。

class Base {...}; 
class ChildA: public Base {...}; 
class ChildB: public Base {...}; 
class GrandChild: public ChildA { 
public: 
    operator ChildB() {...} 
}; 

__global__ void mykernel(Base x) {...} 
__global__ void mykernel2(ChildB y) {...} 

int main() { 
    GrandChild x; 
    mykernel<<<1,1>>>(x); 
    mykernel2<<<1,1>>>(ChildB(x)); 
    return 0; 
}