Invio di array 2D a Cuda Kernel

Question

Ho un po 'di problemi a capire come inviare un array 2D a Cuda. Ho un programma che analizza un file di grandi dimensioni con 30 punti dati su ogni riga. Ho letto circa 10 righe alla volta e poi ho creato una matrice per ogni riga e ogni elemento (quindi nel mio esempio di 10 righe con 30 punti dati, sarebbe int list[10][30]; Il mio obiettivo è quello di inviare questo array al mio kernal e avere ogni processo di blocco una riga (ho ottenuto questo per funzionare perfettamente in C normale, ma Cuda è stato un po 'più impegnativo)

Ecco cosa sto facendo finora ma senza fortuna (nota: sizeofbucket = rows e sizeOfBucketsHoldings = items in fila ... so che dovrei vincere un premio per i nomi delle variabili dispari):

int list[sizeOfBuckets][sizeOfBucketsHoldings]; //this is created at the start of the file and I can confirmed its filled with the correct data 
#define sizeOfBuckets 10 //size of buckets before sending to process list 
#define sizeOfBucketsHoldings 30 
    //Cuda part 
       //define device variables 
       int *dev_current_list[sizeOfBuckets][sizeOfBucketsHoldings]; 
       //time to malloc the 2D array on device 
       size_t pitch; 
       cudaMallocPitch((int**)&dev_current_list, (size_t *)&pitch, sizeOfBucketsHoldings * sizeof(int), sizeOfBuckets); 

       //copy data from host to device 
       cudaMemcpy2D(dev_current_list, pitch, list, sizeOfBuckets * sizeof(int), sizeOfBuckets * sizeof(int), sizeOfBucketsHoldings * sizeof(int),cudaMemcpyHostToDevice); 

       process_list<<<count,1>>> (sizeOfBuckets, sizeOfBucketsHoldings, dev_current_list, pitch); 
       //free memory of device 
       cudaFree(dev_current_list); 


    __global__ void process_list(int sizeOfBuckets, int sizeOfBucketsHoldings, int *current_list, int pitch) { 
     int tid = blockIdx.x; 
     for (int r = 0; r < sizeOfBuckets; ++r) { 
      int* row = (int*)((char*)current_list + r * pitch); 
      for (int c = 0; c < sizeOfBucketsHoldings; ++c) { 
       int element = row[c]; 
      } 
     } 

L'errore che sto ottenendo è:

main.cu(266): error: argument of type "int *(*)[30]" is incompatible with parameter of type "int *" 
1 error detected in the compilation of "/tmp/tmpxft_00003f32_00000000-4_main.cpp1.ii". 

linea 266 è la chiamata del kernel process_list<<<count,1>>> (count, countListItem, dev_current_list, pitch); Penso che il problema è che sto cercando di creare il mio array nella mia funzione di int * ma in quale altro modo posso crearlo? Nel mio codice C puro, io uso int current_list[num_of_rows][num_items_in_row] che funziona ma non riesco a ottenere lo stesso risultato per funzionare in Cuda.

Il mio obiettivo finale è semplice Voglio solo ottenere ogni blocco per elaborare ogni riga (sizeOfBuckets) e quindi farlo scorrere attraverso tutti gli elementi in quella riga (sizeOfBucketHoldings). Inizialmente ho appena fatto un normale cudamalloc e cudaMemcpy ma non funzionava, quindi mi sono guardato intorno e ho scoperto MallocPitch e 2dcopy (entrambi non erano nel mio libro cuda by example) e ho cercato di studiare degli esempi, ma sembra che siano dandomi lo stesso errore (sto leggendo la guida alla programmazione CUDA_C ho trovato questa idea a pagina 22 ma ancora senza fortuna). Qualche idea? o suggerimenti su dove cercare?

Modifica: Per verificare questo, voglio solo aggiungere il valore di ogni riga insieme (ho copiato la logica dal cuda esempio esempio esempio di array). mio kernel:

__global__ void process_list(int sizeOfBuckets, int sizeOfBucketsHoldings, int *current_list, size_t pitch, int *total) { 
    //TODO: we need to flip the list as well 
    int tid = blockIdx.x; 
    for (int c = 0; c < sizeOfBucketsHoldings; ++c) { 
     total[tid] = total + current_list[tid][c]; 
    } 
} 

Ecco come dichiaro la matrice totale nel mio principale:

int *dev_total; 
cudaMalloc((void**)&dev_total, sizeOfBuckets * sizeof(int)); 

Answer 1

Hai alcuni errori nel codice.

• Quindi si copia l'array host sul dispositivo, si dovrebbe passare un puntatore host unidimensionale. Vedere lo function signature.
• Non è necessario allocare array 2D statici per la memoria del dispositivo. Crea array statici nella memoria dell'host, quindi lo si ricrea come array di dispositivi. Tieni presente che deve essere un array monodimensionale. Vedi questo function signature.

Questo esempio dovrebbe aiutare con allocazione di memoria:

__global__ void process_list(int sizeOfBucketsHoldings, int* total, int* current_list, int pitch) 
{ 
    int tid = blockIdx.x; 
    total[tid] = 0; 
    for (int c = 0; c < sizeOfBucketsHoldings; ++c) 
    { 
     total[tid] += *((int*)((char*)current_list + tid * pitch) + c); 
    } 
} 

int main() 
{ 
    size_t sizeOfBuckets   = 10; 
    size_t sizeOfBucketsHoldings = 30; 

    size_t width = sizeOfBucketsHoldings * sizeof(int);//ned to be in bytes 
    size_t height = sizeOfBuckets; 

    int* list = new int [sizeOfBuckets * sizeOfBucketsHoldings];// one dimensional 
    for (int i = 0; i < sizeOfBuckets; i++) 
     for (int j = 0; j < sizeOfBucketsHoldings; j++) 
      list[i *sizeOfBucketsHoldings + j] = i; 

    size_t pitch_h = sizeOfBucketsHoldings * sizeof(int);// always in bytes 

    int* dev_current_list; 
    size_t pitch_d; 
    cudaMallocPitch((int**)&dev_current_list, &pitch_d, width, height); 

    int *test; 
    cudaMalloc((void**)&test, sizeOfBuckets * sizeof(int)); 
    int* h_test = new int[sizeOfBuckets]; 

    cudaMemcpy2D(dev_current_list, pitch_d, list, pitch_h, width, height, cudaMemcpyHostToDevice); 

    process_list<<<10, 1>>>(sizeOfBucketsHoldings, test, dev_current_list, pitch_d); 
    cudaDeviceSynchronize(); 

    cudaMemcpy(h_test, test, sizeOfBuckets * sizeof(int), cudaMemcpyDeviceToHost); 

    for (int i = 0; i < sizeOfBuckets; i++) 
     printf("%d %d\n", i , h_test[i]); 
    return 0; 
} 

Per accedere l'array 2D in kernel si dovrebbe usare modello base_addr + y * pitch_d + x. ATTENZIONE: il pitvh sempre in byte. Devi lanciare il puntatore su byte*.