Creare dinamiche di array locali all'interno del kernel OpenCL

Question

Ho un kernel OpenCL che deve elaborare una matrice come array multipli in cui ogni somma di sub-array viene salvata in un array di cache locale.

Ad esempio, immaginare la matrice fowling:

[[1, 2, 3, 4], [10, 30, 1, 23]] 

• Ciascun gruppo di lavoro ottiene una matrice (nel exemple abbiamo 2 gruppi di lavoro);

• Ciascun elemento di lavoro elabora due indici di array (ad esempio, moltiplica l'indice di valore id_locale), in cui il risultato dell'elemento di lavoro viene salvato in un array condiviso del gruppo di lavoro.

  __kernel void test(__global int **values, __global int *result, const int array_size){ 
      __local int cache[array_size]; 
  
      // initialise 
      if (get_local_id(0) == 0){ 
       for (int i = 0; i < array_size; i++) 
        cache[i] = 0; 
      } 
  
      barrier (CLK_LOCAL_MEM_FENCE); 
  
      if(get_global_id(0) < 4){ 
       for (int i = 0; i<2; i++) 
        cache[get_local_id(0)] += values[get_group_id(0)][i] * 
                   get_local_id(0); 
      } 
  
      barrier (CLK_LOCAL_MEM_FENCE); 
  
      if(get_local_id(0) == 0){ 
       for (int i = 0; i<array_size; i++) 
        result[get_group_id(0)] += cache[i]; 
      } 
  } 
  

Il problema è che non riesco a definire la dimensione dell'array cache utilizzando un parametro del kernel, ma devo per avere un kernel dinamico.

Come posso creare dinamicamente? come la funzione malloc in c ...

Oppure l'unica soluzione disponibile è quella di inviare un array temporaneo alla mia funzione del kernel?

Answer 1

Ciò può essere ottenuto aggiungendo __local array come parametro del kernel:

__kernel void test(__global int **values, __global int *result, 
    const int array_size, __local int * cache) 

e fornendo dimensione desiderata del parametro del kernel:

clSetKernelArg(kernel, 3, array_size*sizeof(int), NULL); 

La memoria locale sarà assegnata l'invocazione kernel . Si noti che potrebbero essere necessari controlli aggiuntivi per garantire che la dimensione della memoria locale richiesta non superi il limite del dispositivo.