從內核調用內核

A跟進Q從：CUDA: Calling a __device__ function from a kernel 從內核調用內核

我試圖加快排序操作。一個簡化版本的僞如下：

// some costly swap operation 
__device__ swap(float* ptrA, float* ptrB){ 
    float saveData;   // swap some 
    saveData= *Adata;  // big complex 
    *Adata= *Bdata   //  data chunk 
    *Bdata= saveData; 
} 

// a rather simple sort operation 
__global__ sort(float data[]){ 
    for (i=0; i<limit: i++){ 
    find left swap point 
    find right swap point 
    swap<<<1,1>>>(left, right); 
    } 
}

（注：這個簡單的版本不顯示在塊還原技術）的想法是，它很容易（快），以確定交換點。交換操作成本很高（很慢）。因此，使用一個塊來查找/識別交換點。使用其他塊進行交換操作。即並行地進行實際的交換。這聽起來像一個體面的計劃。但是，如果編譯器在設備調用中插入行，那麼就不會發生並行交換。有沒有辦法告訴編譯器不要內聯設備調用？

來源

2012-07-31 Doug

編輯（2016）：

動態並行是在第二代開普勒架構GPU的引入。計算能力3.5及更高版本的設備支持在設備中啓動內核。

原來的答案：

你將不得不等待，直到今年結束的時候，新一代的硬件是可用的。目前沒有任何CUDA設備可以從其他內核啓動內核 - 目前它不受支持。

來源

2012-07-31 19:15:38 talonmies

是否可以在最新版本的CUDA（v6.5）和具有計算能力3.0的NVIDIA Grid K520上執行此操作？ – 2015-03-25 19:09:49

@talonmies，我已經等了4年你的答案：D希望有這樣的解決方案:-)你能指出我嗎？ – Nabin 2016-03-11 05:41:05

沒關係。我找到了什麼 – Nabin 2016-03-11 05:56:59

我知道問這個問題已經很久了。當我搜索同樣的問題時，我到了這個頁面。似乎我得到了解決方案。

解決方案：

我伸手here莫名其妙，看到清涼的做法從另一個內核中啓動內核。

__global__ void kernel_child(float *var1, int N){ 
    //do data operations here 
} 


__global__ void kernel_parent(float *var1, int N) 
{ 
    kernel_child<<<1,2>>>(var1,N); 
}

cuda 5.0及以上的動態並行性使這成爲可能。同時在運行時確保使用compute_35架構或更高版本。

終端方式 您可以從終端運行上面的父內核（它將最終運行子內核）。在Linux機器上驗證。

$ nvcc -arch=sm_35 -rdc=true yourFile.cu 
$ ./a.out

希望它有幫助。謝謝！

來源

2016-03-11 06:02:51 Nabin

從內核調用內核

回答

相關問題