2011-03-08 24 views
15

He pasado por muchos foros y nvidia manual, pero no pude entender lo que es __threadfence() y el uso de él?CUDA __threadfence()

Gracias.

Respuesta

40

Normalmente, no hay garantía de que si un bloque escribe algo en la memoria global, el otro bloque "lo verá". Tampoco se garantiza el orden de las escrituras en la memoria global, con la excepción del bloque que lo emitió.

Hay dos excepciones:

  • operaciones atómicas - los que siempre son visibles por otros bloques
  • threadfence

Imagínese, que un bloque produce algunos datos, y luego usa operación atómica de marque una bandera que los datos están allí. Pero es posible que el otro bloque vea la bandera, pero leerá datos incorrectos o incompletos.

__threadfence función detiene el hilo actual hasta que sus escrituras a la memoria global se garanticen para ser visibles por todos los otros hilos en la red. Por lo tanto, si haces algo como:

  1. almacenar sus datos
  2. __threadfence()
  3. marcan atómicamente una bandera

se garantiza que si el otro bloque ve la bandera, también verá los datos.

Más información: Guía de programación Cuda, capítulos B.2.4 y B.5

+0

¿Qué tal __syncthreads(), hace que la garantía de que los accesos a memoria por cualquier tema en el bloque será visible a todos los hilos en el bloque? –

+2

'__syncthreads()' es más fuerte que '__threadfence_block()'. Después de '__syncthreads()' tiene la garantía de que todas las escrituras de memoria compartida/global antes de la barrera son visibles por todos los hilos después de la barrera. Sin embargo, '__syncthreads()' tiene un efecto solo en un bloque y no se dan garantías entre subprocesos de diferentes bloques. – CygnusX1

+0

L1 de otros SMs será coherente con los datos almacenados? ¿O todavía necesita especificar cargas de alcance global (L1-no-cacheable)? – maxbc