Quand est-CUDA de l' __shared__ la mémoire est utile?

Dans le cas que vous mentionnez, la mémoire partagée n'est pas utile, pour la raison suivante: chaque élément de données est utilisé qu'une seule fois. Pour la mémoire partagée pour etre utile, vous devez utiliser les données transférées vers la mémoire partagée à plusieurs reprises, en utilisant les bons modèles d'accès, pour l'avoir aider. La raison pour cela est simple: il suffit de lecture de la mémoire globale nécessite 1 mondial de lire la mémoire et le zéro de la mémoire partagée lit; la lecture en mémoire partagée première demanderait 1 mondial de lecture de la mémoire et 1 mémoire partagée lecture, qui prend plus de temps.

Voici un exemple simple, où chaque fil dans le bloc calcule la valeur correspondante, au carré, plus la moyenne des deux sa gauche et droit voisins, au carré:

  __global__ void compute_it(float *data)
  {
     int tid = threadIdx.x;
     __shared__ float myblock[1024];
     float tmp;

     //load the thread's data element into shared memory
     myblock[tid] = data[tid];

     //ensure that all threads have loaded their values into
     //shared memory; otherwise, one thread might be computing
     //on unitialized data.
     __syncthreads();

     //compute the average of this thread's left and right neighbors
     tmp = (myblock[tid > 0 ? tid - 1 : 1023] + myblock[tid < 1023 ? tid + 1 : 0]) * 0.5f;
     //square the previousr result and add my value, squared
     tmp = tmp*tmp + myblock[tid] * myblock[tid];

     //write the result back to global memory
     data[tid] = tmp;
  }

Noter que ce qui est envisagé, travailler à l'aide d'un seul bloc. L'extension à plus de blocs doit être simple. Suppose bloc de dimension (1024, 1, 1) et de la grille de dimension (1, 1, 1).