De l'accès non coalesced à l'accès mémoire coalesced CUDA

Question

Je me demandais s'il existe un moyen simple de transformer un accès mémoire non coalesced en un coalesced. Prenons l'exemple de ce tableau:

dW[[w0,w1,w2][w3,w4,w5][w6,w7][w8,w9]] 

Maintenant, je sais que si cette discussion 0 dans le bloc 0 accès dW[0] puis enfilez 1 dans le bloc 0 accès dw[1], c'est un accès coalisée dans la mémoire globale. Le problème est que j'ai deux opérations. Le premier est coalesced comme décrit ci-dessus. Mais le second n'est pas parce que Thread 1 dans le bloc 0 doit effectuer une opération à la fois sur dW[0], dW[1] et dW[2].

Je sais que la forme initiale du conteneur permet ou interdit l'accès coalescé. Mais dW est un très grand tableau, et je ne peux pas le transformer pendant le processus.

Savez-vous s'il est possible d'atténuer ce problème?

Answer 1

Vous pouvez essayer d'utiliser la mémoire partagée peut-être, cela pourrait fonctionner (ou non, difficile à dire sans un exemple). Par exemple, disons que le premier accès à l'opération a coalescé les données et que le second a beaucoup progressé; cela peut speedup choses

__shared__ int shared[BLOCK_SIZE]; 
// Load data global -> shared with coalesced access ; you may need to load a bit more before/after depending on you application 
shared[tid] = global[some id] 
syncthreads(); 
// Do the math with coalescing access 
function0(shared[tid]) 
// Do the math with the non coalescing access 
function1(shared[tid+-1 or wathever]) 

L'idée est de charger des données en commun d'une manière coalescent, puis utilisez partagé pour faire le calcul, car l'accès coalescent n'importe avec mémoire partagée (mais le conflit bancaire faire de l'autre main, c'est généralement bien quand même).

Vous devrez nous donner plus d'informations si vous voulez une aide plus précise. C'est juste un indice.