2012-04-21 18 views
12

De acuerdo con "CUDA C Programming Guide", un acceso a memoria constante se beneficia solo si se golpea un caché de multiprocesador constante (Sección 5.3.2.4) . De lo contrario, puede haber incluso más solicitudes de memoria para una mitad de urdimbre que en el caso de la lectura de memoria global fusionada. Entonces, ¿por qué el tamaño de memoria constante está limitado a 64 KB?¿Por qué el tamaño de memoria constante es limitado en CUDA?

Una pregunta más para no preguntar dos veces. Por lo que yo entiendo, en la arquitectura de Fermi, la caché de texturas se combina con la caché L2. ¿El uso de texturas sigue teniendo sentido o las lecturas de la memoria global se almacenan en la memoria caché de la misma manera?


memoria constante (sección 5.3.2.4)

El espacio de memoria constante reside en la memoria del dispositivo y se almacena en caché en la memoria caché constante mencionado en las Secciones F.3.1 y F. 4.1.

Para dispositivos de capacidad informática 1.x, una solicitud de memoria constante para un warp primero se divide en dos solicitudes, una para cada half-warp, que se emiten de forma independiente.

Una solicitud se divide en tantas solicitudes separadas como diferentes direcciones de memoria en la solicitud inicial, disminuyendo el rendimiento en un factor igual al número de solicitudes por separado.

Las solicitudes resultantes se revisan con el rendimiento de la memoria caché constante en caso de un golpe de caché, o en el rendimiento de la memoria del dispositivo en caso contrario.

Respuesta

15

El tamaño de la memoria constante es de 64 KB para la capacidad de cálculo de 1.0-3.0 dispositivos. El conjunto de trabajo de caché es de solo 8 KB (consulte la Guía de programación de CUDA v4.2, tabla F-2).

La memoria constante es utilizada por el controlador, el compilador y las variables declaradas __device__ __constant__. El controlador usa memoria constante para comunicar parámetros, enlaces de textura, etc. El compilador usa constantes en muchas de las instrucciones (ver desmontaje).

Las variables colocadas en la memoria constante se pueden leer y escribir utilizando las funciones de tiempo de ejecución del host cudaMemcpyToSymbol() y cudaMemcpyFromSymbol() (consulte la Guía de programación de CUDA v4.2 sección B.2.2). La memoria constante está en la memoria del dispositivo, pero se accede a través de la memoria caché constante.

En la textura Fermi, constante, L1 y I-Cache son cachés de nivel 1 en o alrededor de cada SM. Todas las memorias caché de nivel 1 acceden a la memoria del dispositivo a través de la memoria caché L2.

El límite constante de 64 KB es por módulo CU que es una unidad de compilación CUDA. El concepto de CUmodule está oculto bajo el tiempo de ejecución de CUDA pero es accesible por la API de controladores de CUDA.

+1

Greg, lo siento por no ser lo suficientemente claro. Sé cómo usar la memoria constante. En la pregunta, me pregunto por qué el tamaño de la memoria constante se limita a 64 KB si solo 8 KB en caché en realidad proporcionan el rendimiento mejor que el de la memoria global. De la misma manera, se puede acceder a la memoria global a través de la memoria caché L1. Entonces, desde el punto de vista del programador, ¿cuál es la diferencia entre usar la memoria global o la memoria constante ya que ambos se almacenan en caché de la misma manera? – AdelNick

+6

La memoria caché constante tiene un tamaño óptimo para gráficos y sombreadores de cómputo. La API de CUDA expone el caché constante para que los desarrolladores puedan aprovechar el caché adicional. La memoria caché constante tiene un rendimiento óptimo cuando todos los hilos acceden a la misma dirección.Los éxitos son muy rápidos. Las señoritas pueden tener el mismo rendimiento que una señorita L1. La memoria caché constante se puede usar para reducir la vibración de la L1. Los dispositivos de capacidad informática 1.x no tienen un caché L1 o L2, por lo que se utilizó el caché constante para optimizar algunos accesos. –

Cuestiones relacionadas