如何在多個內核和多個GPU中共享全局內存？

Question

----------------a.c--------------------- 
variable *XX; 
func1(){ 
    for(...){ 
    for(i = 0; i < 4; i++) 
     cutStartThread(func2,args) 
    } 
} 
---------------b.cu------------------- 
func2(args){ 
    cudaSetDevice(i); 
    xx = cudaMalloc(); 
    mykernel<<<...>>>(xx); 
} 
-------------------------------------- 

最近，我想爲我的程序使用多個GPU設備。我的節點上有四個Tesla C2075卡。我使用四個線程來管理四個GPU。更重要的是，每個線程中的內核都會多次啓動。一個簡單的僞代碼如上。我有兩個問題：

1. 可變XX是一個很長的字符串，只有在內核中讀取。我想在mykernel的多次發佈期間保留它。只有在mykernel第一次啓動時纔可以致電cudaMalloc並將指針傳遞給mykernel？或者我應該使用__device__限定符？

2. XX在四個線程中使用，所以我將其聲明爲文件a.c中的全局變量。 XX的多個cudaMalloc是否正確，還是應該使用數組，例如variable *xx[4]？

Answer 1

1. 對於由一個單一的設備上運行的內核使用情況，您可以撥打cudaMalloc一次以創建您的變量XX拿着字符串，然後路過cudaMalloc（即XX）創建的指針爲準內核需要它。

   #define xx_length 20 
   char *XX; 
   cudaMalloc((void **)&XX, xx_length * sizeof(char)); 
   ... 
   kernel1<<<...>>>(XX, ...); 
   ... 
   kernel2<<<...>>>(XX, ...); 
   etc. 
   

2. 創建針對每個線程獨立的XX變量，假設每個線程正被用於訪問一個不同的設備。你如何做到這一點將取決於XX的範圍。但是數組：在全局範圍內

   char *XX[num_devices]; 
   

，應該沒問題。

CUDA OpenMP sample作爲如何使用多個線程來管理多個GPU的例子可能是有意義的。