[Optimization] Efficienza delle dimensioni di blocchi e griglie CUDA


Answers

Ok suppongo che abbiamo a che fare con due domande qui.

1) Un buon modo per assegnare dimensioni del blocco (cioè il numero di thread) Questo di solito dipende dal tipo di dati che stai trattando. Hai a che fare con i vettori? Hai a che fare con le matrici? Il modo suggerito è di mantenere il numero di thread in multipli di 32. Quindi, quando si tratta di vettori, l'avvio di 256 x 1, 512 x 1 blocchi può andare bene. E similariy quando si tratta di matrici, 32 x 8, 32 x 16.

2) Un buon modo per assegnare le dimensioni della griglia (ovvero il numero di blocchi) Qui diventa un po 'complicato. È sufficiente lanciare 10.000 blocchi perché non è normalmente il modo migliore di fare le cose. Cambiare i blocchi dentro e fuori l'hardware è costoso. Due cose da considerare sono la memoria condivisa utilizzata per blocco e il numero totale di SP disponibili e risolvere il numero ottimale.

Puoi trovare una buona implementazione di come farlo dalla spinta . Tuttavia, potrebbe volerci un po 'per capire cosa sta succedendo all'interno del codice.

Question

Qual è il modo consigliato di trattare i dataset di dimensioni dinamiche in cuda?

Si tratta di "impostare il blocco e le dimensioni della griglia in base al set di problemi" o vale la pena assegnare le dimensioni dei blocchi come fattori di 2 e disporre di una logica in-kernel per gestire l'over-spill?

Posso vedere come questo probabilmente importi molto per le dimensioni del blocco, ma quanto conta questo per le dimensioni della griglia? A mio avviso, i vincoli hardware effettivi si fermano a livello di blocco (cioè i blocchi assegnati a SM che hanno un determinato numero di SP, e quindi in grado di gestire una particolare dimensione di curvatura).

Ho esaminato la "Programmazione di processori paralleli in parallelo" di Kirk, ma in realtà non tocca proprio quest'area.




Se si dispone di set di dati dimensionati dinamicamente, è probabile che si verifichino alcuni problemi con la latenza mentre alcuni thread e blocchi attendono il completamento degli altri.

Questo sito ha delle ottime euristiche. Alcuni punti salienti generali:

Scelta dei blocchi per griglia

  • I blocchi per griglia dovrebbero essere> = numero di multiprocessori.
  • __syncthreads() uso di __syncthreads() nei kernel, più blocchi (in modo che un blocco possa essere eseguito mentre un altro attende la sincronizzazione)

Scelta di thread per blocco

  • Discussioni in multipli di dimensione di curvatura (cioè in genere 32)

  • Generalmente utile scegliere il numero di thread in modo che il numero massimo di thread per blocco (basato su hardware) sia un multiplo del numero di thread. Ad esempio con thread max di 768, l'utilizzo di 256 thread per blocco tenderà a essere migliore di 512 perché più thread possono essere eseguiti contemporaneamente su un blocco.