¿Qué tipo de problemas se prestan bien a la informática de GPU?

84

Así que tengo una cabeza decente con respecto a qué problemas con los que trabajo son los mejores en serie, y cuáles se pueden manejar en paralelo. Pero en este momento, no tengo mucha idea de lo que se maneja mejor con el cómputo basado en la CPU, y qué se debe descargar a una GPU.

Sé que es una pregunta básica, pero gran parte de mi búsqueda queda atrapada en personas que claramente defienden uno u otro sin justificar realmente por qué , o reglas generales algo vagas. Buscando una respuesta más útil aquí.

Fomite
fuente

Respuestas:

63

El hardware de la GPU tiene dos fortalezas particulares: cómputo sin procesar (FLOP) y ancho de banda de memoria. Los problemas computacionales más difíciles caen en una de estas dos categorías. Por ejemplo, el álgebra lineal densa (A * B = C o Solve [Ax = y] o Diagonalize [A], etc.) cae en algún lugar del espectro de ancho de banda de cálculo / memoria dependiendo del tamaño del sistema. Las transformaciones rápidas de Fourier (FFT) también se ajustan a este molde con altas necesidades de ancho de banda agregado. Al igual que otras transformaciones, algoritmos basados ​​en cuadrícula / malla, Monte Carlo, etc. Si observa los ejemplos de código del SDK de NVIDIA , puede tener una idea de los tipos de problemas que se abordan con mayor frecuencia.

Creo que la respuesta más instructiva es la pregunta "¿En qué tipo de problemas son realmente malas las GPU?" La mayoría de los problemas que no entran en esta categoría pueden ejecutarse en la GPU, aunque algunos requieren más esfuerzo que otros.

Los problemas que no se mapean bien son generalmente demasiado pequeños o demasiado impredecibles. Los problemas muy pequeños carecen del paralelismo necesario para usar todos los hilos en la GPU y / o podrían caber en un caché de bajo nivel en la CPU, lo que aumenta sustancialmente el rendimiento de la CPU. Los problemas impredecibles tienen demasiadas ramas significativas, lo que puede evitar que los datos se transmitan de manera eficiente desde la memoria de la GPU a los núcleos o reducir el paralelismo al romper el paradigma SIMD (consulte ' deformaciones divergentes '). Los ejemplos de este tipo de problemas incluyen:

  • La mayoría de los algoritmos gráficos (demasiado impredecibles, especialmente en el espacio de memoria)
  • Álgebra lineal escasa (pero esto también es malo en la CPU)
  • Pequeños problemas de procesamiento de señal (FFTs menores de 1000 puntos, por ejemplo)
  • Buscar
  • Ordenar
Max Hutchinson
fuente
3
Aún así, las soluciones de GPU para esos problemas "impredecibles" son posibles y, aunque hoy en día no son factibles, pueden tener importancia en el futuro.
Leftaroundabout
66
Me gustaría agregar ramas específicamente a la lista de interruptores de rendimiento de GPU. Desea que todos sus (cientos) ejecuten la misma instrucción (como en SIMD) para realizar un cálculo verdaderamente paralelo. Por ejemplo, en tarjetas AMD si alguno de los flujos de instrucciones encuentra una rama y debe divergir, todo el frente de onda (grupo paralelo) diverge. Si otras unidades del frente de onda no deben divergir, deben realizar un segundo pase. Supongo que eso es lo que Maxhutch quiere decir con previsibilidad.
Violet Giraffe
2
@VioletGiraffe, eso no es necesariamente cierto. En CUDA (es decir, en las GPU Nvidia), la divergencia de ramificación solo afecta a la deformación actual, que es de 32 hilos como máximo. Distorsiones diferentes, aunque ejecutan el mismo código, no son síncronas a menos que estén explícitamente sincronizadas (por ejemplo, con __synchtreads()).
Pedro
1
@Pedro: Cierto, pero la ramificación en general perjudica el rendimiento. Para los códigos de alto rendimiento (¿qué no es el código GPU?), Es casi esencial tenerlo en cuenta.
jvriesem
21

Los problemas que tienen una alta intensidad aritmética y patrones regulares de acceso a la memoria suelen ser fáciles de implementar en las GPU y funcionan bien en ellos.

La dificultad básica para tener un código GPU de alto rendimiento es que tiene una tonelada de núcleos y desea que se utilicen todos en su máxima potencia tanto como sea posible. Los problemas que tienen patrones irregulares de acceso a la memoria o que no tienen una intensidad aritmética alta hacen que esto sea difícil: o pasas mucho tiempo comunicando resultados o pasas mucho tiempo buscando cosas de la memoria (¡lo cual es lento!), Y no tienes suficiente tiempo para hacer números. Por supuesto, el potencial de concurrencia en su código es crítico para su capacidad de implementarse bien en la GPU también.

Reid.Atcheson
fuente
¿Puede especificar qué quiere decir con patrones de acceso a memoria regulares?
Fomite el
1
La respuesta de Maxhutch es mejor que la mía. Lo que quiero decir con patrón de acceso regular es que se accede a la memoria de manera temporal y espacialmente local. Es decir: no haces saltos enormes alrededor de la memoria repetidamente. También es algo así como un paquete que he notado. También se entiende que sus patrones de acceso a datos pueden ser predeterminados por el compilador de alguna manera o por el programador, de modo que se minimice la ramificación (declaraciones condicionales en el código).
Reid.Atcheson el
15

Esto no pretende ser una respuesta en sí misma, sino más bien una adición a las otras respuestas de maxhutch y Reid.Atcheson .

Para sacar el máximo provecho de las GPU, su problema no solo necesita ser altamente (o masivamente) paralelo, sino que también el algoritmo central que se ejecutará en la GPU debe ser lo más pequeño posible. En términos de OpenCL , esto se conoce principalmente como el núcleo .

Para ser más precisos, el núcleo debe caber en el registro de cada unidad de multiprocesamiento (o unidad de cálculo ) de la GPU. El tamaño exacto del registro depende de la GPU.

Dado que el núcleo es lo suficientemente pequeño, los datos sin procesar del problema deben caber en la memoria local de la GPU (léase: memoria local (OpenCL) o memoria compartida (CUDA) de una unidad de cómputo). De lo contrario, incluso el alto ancho de banda de memoria de la GPU no es lo suficientemente rápido como para mantener ocupados los elementos de procesamiento todo el tiempo.
Por lo general, esta memoria es de aproximadamente 16 a 32 KiByte grande .

Torbjörn
fuente
¿No se comparte la memoria local / compartida de cada unidad de procesamiento entre todas las docenas (?) De subprocesos que se ejecutan dentro de un solo grupo de núcleos? En este caso, ¿no necesita realmente mantener su conjunto de datos de trabajo significativamente más pequeño para obtener el rendimiento completo de la GPU?
Dan Neely
La memoria local / compartida de una unidad de procesamiento solo es accesible por la propia unidad de cálculo y, por lo tanto, solo es compartida por los elementos de procesamiento de esta unidad de cálculo. Todas las unidades de procesamiento pueden acceder a la memoria global de la tarjeta gráfica (generalmente 1 GB). El ancho de banda entre los elementos de procesamiento y la memoria local / compartida es muy rápido (> 1TB / s) pero el ancho de banda a la memoria global es mucho más lento (~ 100GB / s) y debe compartirse entre todas las unidades de cómputo.
Torbjörn
No estaba preguntando sobre la memoria principal de la GPU. Pensé que la memoria en el dado solo se asignaba en el grupo de nivel de núcleo, no por núcleo individual. ex para un nVidia GF100 / 110 gpu; para cada uno de los 16 grupos SM no los 512 núcleos cuda. Con cada SM diseñado para ejecutar hasta 32 hilos en paralelo, maximizar el rendimiento de la GPU requeriría mantener el conjunto de trabajo en el rango de 1 kb / hilo.
Dan Neely
@Torbjoern Lo que desea es mantener ocupadas todas las canalizaciones de ejecución de GPU, las GPU logran esto de dos maneras: (1) la forma más común es aumentar la ocupación, o dicho de otra manera, al aumentar el número de hilos concurrentes (los núcleos pequeños usan menos los recursos compartidos para que pueda tener hilos más activos); quizás mejor, es (2) aumentar el paralelismo del nivel de instrucción dentro de su núcleo, para que pueda tener un núcleo más grande con una ocupación relativamente baja (pequeño número de hilos activos). Ver bit.ly/Q3KdI0
fcruz
11

Probablemente una adición más técnica a las respuestas anteriores: las GPU CUDA (es decir, Nvidia) se pueden describir como un conjunto de procesadores que funcionan de forma autónoma en 32 subprocesos cada uno. Los subprocesos en cada procesador funcionan en paso de bloqueo (piense en SIMD con vectores de longitud 32).

Aunque la forma más tentadora de trabajar con GPU es pretender que absolutamente todo funciona en un paso de bloqueo, esta no siempre es la forma más eficiente de hacer las cosas.

Si su código no paralelizar muy bien / automáticamente a cientos / miles de hilos, que pueden ser capaces de descomponerlo en tareas asíncronas individuales que no paralelizan bien, y ejecutan aquellos con sólo 32 subprocesos que se ejecutan en el bloqueo de paso. CUDA proporciona un conjunto de instrucciones atómicas que permiten implementar mutexes que a su vez permite que los procesadores se sincronicen entre sí y procesen una lista de tareas en un paradigma de grupo de subprocesos . Su código funcionaría de la misma manera que en un sistema multinúcleo, solo tenga en cuenta que cada núcleo tiene 32 hilos propios.

Aquí hay un pequeño ejemplo, usando CUDA, de cómo funciona esto

/* Global index of the next available task, assume this has been set to
   zero before spawning the kernel. */
__device__ int next_task;

/* We will use this value as our mutex variable. Assume it has been set to
   zero before spawning the kernel. */
__device__ int tasks_mutex;

/* Mutex routines using atomic compare-and-set. */
__device__ inline void cuda_mutex_lock ( int *m ) {
    while ( atomicCAS( m , 0 , 1 ) != 0 );
    }
__device__ inline void cuda_mutex_unlock ( int *m ) {
    atomicExch( m , 0 );
    }

__device__ void task_do ( struct task *t ) {

    /* Do whatever needs to be done for the task t using the 32 threads of
       a single warp. */
    }

__global__ void main ( struct task *tasks , int nr_tasks ) {

    __shared__ task_id;

    /* Main task loop... */
    while ( next_task < nr_tasks ) {

        /* The first thread in this block is responsible for picking-up a task. */
        if ( threadIdx.x == 0 ) {

            /* Get a hold of the task mutex. */
            cuda_mutex_lock( &tasks_mutex );

            /* Store the next task in the shared task_id variable so that all
               threads in this warp can see it. */
            task_id = next_task;

            /* Increase the task counter. */
            next_tast += 1;

            /* Make sure those last two writes to local and global memory can
               be seen by everybody. */
            __threadfence();

            /* Unlock the task mutex. */
            cuda_mutex_unlock( &tasks_mutex );

            }

        /* As of here, all threads in this warp are back in sync, so if we
           got a valid task, perform it. */
        if ( task_id < nr_tasks )
            task_do( &tasks[ task_id ] );

        } /* main loop. */

    }

Luego debe llamar al núcleo main<<<N,32>>>(tasks,nr_tasks)para asegurarse de que cada bloque contenga solo 32 subprocesos y, por lo tanto, quepa en una sola urdimbre. En este ejemplo, también asumí, por simplicidad, que las tareas no tienen dependencias (por ejemplo, una tarea depende de los resultados de otra) o conflictos (por ejemplo, el trabajo en la misma memoria global). Si este es el caso, la selección de tareas se vuelve un poco más complicada, pero la estructura es esencialmente la misma.

Esto es, por supuesto, más complicado que simplemente hacer todo en un gran lote de celdas, pero amplía significativamente el tipo de problemas para los que se pueden usar las GPU.

Pedro
fuente
2
Esto es técnicamente cierto, pero se necesita un alto paralelismo para obtener un ancho de banda de memoria alto y hay un límite para la cantidad de llamadas asíncronas del núcleo (actualmente 16). También hay toneladas de comportamiento indocumentado relacionado con la programación en la versión actual. Aconsejaría no depender de núcleos asíncronos para mejorar el rendimiento por el momento ...
Max Hutchinson
2
Lo que estoy describiendo se puede hacer todo en una sola llamada del núcleo. Puede hacer N bloques de 32 hilos cada uno, de modo que cada bloque encaje en una sola urdimbre. Luego, cada bloque adquiere una tarea de una lista global de tareas (acceso controlado mediante atómicos / mutexes) y la calcula utilizando 32 subprocesos bloqueados. Todo esto sucede en una sola llamada del núcleo. Si desea un ejemplo de código, avíseme y lo publicaré.
Pedro
4

Un punto que no se ha hecho hasta ahora es que la generación actual de GPU no funciona tan bien en los cálculos de coma flotante de doble precisión como en los cálculos de precisión simple. Si sus cálculos tienen que hacerse con doble precisión, entonces puede esperar que el tiempo de ejecución aumente en un factor de aproximadamente 10 sobre la precisión simple.

Brian Borchers
fuente
Quiero estar en desacuerdo La mayoría (o todas) las GPU más nuevas tienen soporte nativo de doble precisión. Casi todas estas GPU informan cálculos de doble precisión que se ejecutan a aproximadamente la mitad de la velocidad de precisión simple, probablemente debido a la simple duplicación de los accesos de memoria / ancho de banda requeridos.
Godric Seer
1
Si bien es cierto que las últimas y mejores tarjetas Nvidia Tesla ofrecen un rendimiento máximo de doble precisión que es la mitad del rendimiento máximo de precisión simple, la proporción es de 8 a 1 para las tarjetas de grado de consumo de arquitectura Fermi más comunes.
Brian Borchers
@GodricSeer La relación 2: 1 de punto flotante SP y DP tiene muy poco que ver con el ancho de banda y casi todo con la cantidad de unidades de hardware que existen para ejecutar estas operaciones. Es común reutilizar el archivo de registro para SP y DP, por lo tanto, la unidad de punto flotante puede ejecutar 2 veces las operaciones SP como operaciones DP. Existen numerosas excepciones a este diseño, por ejemplo, IBM Blue Gene / Q (no tiene lógica SP y, por lo tanto, SP se ejecuta a ~ 1.05x DP). Algunas GPU tienen proporciones distintas de 2, por ejemplo, 3 y 5.
Jeff
Han pasado cuatro años desde que escribí esta respuesta, y la situación actual con las GPU NVIDIA es que para las líneas GeForce y Quadro, la relación DP / SP es ahora 1/32. Las GPU Tesla de NVIDIA tienen un rendimiento de doble precisión mucho más fuerte, pero también cuestan mucho más. Por otro lado, AMD no ha paralizado el rendimiento de doble precisión en sus GPU Radeon de la misma manera.
Brian Borchers
4

Desde un punto de vista metafórico, el gpu se puede ver como una persona acostada en una cama de clavos. La persona que está en la parte superior son los datos y en la base de cada uña hay un procesador, por lo que la uña es en realidad una flecha que apunta del procesador a la memoria. Todas las uñas tienen un patrón regular, como una cuadrícula. Si el cuerpo está bien extendido, se siente bien (el rendimiento es bueno), si el cuerpo solo toca algunos puntos del lecho ungueal, entonces el dolor es malo (mal desempeño).

Esto se puede tomar como una respuesta complementaria a las excelentes respuestas anteriores.

labotsirc
fuente
4

Antigua pregunta, pero creo que esta respuesta de 2014 , relacionada con métodos estadísticos, pero generalizable para cualquiera que sepa lo que es un ciclo, es particularmente ilustrativa e informativa.

GT.
fuente
2

Las GPU tienen E / S de latencia prolongada, por lo que se deben utilizar muchos subprocesos para saturar la memoria. Mantener una urdimbre ocupada requiere muchos hilos. Si la ruta del código es de 10 relojes y la latencia de E / S 320 relojes, 32 hilos deberían estar cerca de saturar la urdimbre. Si la ruta del código es de 5 relojes, duplique los hilos.

Con mil núcleos, busque miles de hilos para utilizar completamente la GPU.

El acceso a la memoria es por línea de caché, generalmente 32 bytes. Cargar un byte tiene un costo comparable a 32 bytes. Por lo tanto, combine el almacenamiento para aumentar la localidad de uso.

Hay muchos registros y RAM local para cada deformación, lo que permite compartir vecinos.

Las simulaciones de proximidad de conjuntos grandes deberían optimizarse bien.

Las E / S aleatorias y el subproceso único es una alegría de matar ...

usuario14381
fuente
Esta es una pregunta realmente fascinante; Estoy discutiendo conmigo mismo sobre si es posible (o vale la pena el esfuerzo) 'hacer paralelamente' una tarea razonablemente sencilla (detección de bordes en imágenes aéreas) cuando cada tarea requiere ~ 0.06 segundos pero hay ~ 1.8 millones de tareas para realizar ( por año, durante 6 años de datos: las tareas son definitivamente separables) ... por lo tanto ~ 7,5 días de tiempo de cálculo en un núcleo. Si cada cálculo fue más rápido en una GPU, y el trabajo podría ser paralelo 1-por-nGPUcores [n pequeño], ¿es probable que el tiempo de trabajo pueda caer a ~ 1 hora? Parece improbable.
GT.
0

Imagine un problema que puede resolverse con mucha fuerza bruta, como el vendedor ambulante. Luego imagine que tiene racks de servidores con 8 tarjetas de video spanky cada una, y cada tarjeta tiene 3000 núcleos CUDA.

Simplemente resuelva TODAS las rutas posibles del vendedor y luego ordene por tiempo / distancia / alguna métrica. Seguro que estás tirando casi el 100% de tu trabajo, pero la fuerza bruta es una solución viable a veces.

Criggie
fuente
Tuve acceso a una pequeña granja de 4 servidores de este tipo durante una semana, y en cinco días hice más bloques de.net distribuidos que en los 10 años anteriores.
Criggie
-1

Al estudiar muchas ideas de ingeniería, diría que un gpu es una forma de enfoque de tareas, de gestión de memoria, de cálculo repetible.

Muchas fórmulas pueden ser simples de escribir pero dolorosas de calcular, como en las matemáticas de matriz no se obtiene una sola respuesta sino muchos valores.

Esto es importante en la computación, ya que tan rápido una computadora calcula valores y ejecuta fórmulas, ya que algunas fórmulas no pueden ejecutarse sin todos los valores calculados (por lo tanto, disminuyen la velocidad). Una computadora no sabe muy bien en qué orden ejecutar fórmulas o calcular valores para usar en estos programas. Principalmente, fuerza bruta a través de velocidades rápidas y divide fórmulas en mandriles para calcular, pero muchos programas en estos días requieren estos mandriles calculados en este momento y esperar en preguntas (y preguntas de preguntas y más preguntas de preguntas).

Por ejemplo, en un juego de simulación que debe calcularse primero en colisiones, el daño de la colisión, la posición de los objetos, la nueva velocidad. ¿Cuánto tiempo debería llevar esto? ¿Cómo puede cualquier CPU manejar esta carga? Además, la mayoría de los programas son muy abstractos y requieren más tiempo para manejar datos y no siempre están diseñados para subprocesos múltiples o no son buenas formas en programas abstractos para hacerlo de manera efectiva.

A medida que la CPU mejoró y la gente mejor se volvió descuidada en la programación, también debemos programar para muchos tipos diferentes de computadoras. Una gpu está diseñada para la fuerza bruta a través de muchos cálculos simples al mismo tiempo (sin mencionar la memoria (secundaria / ram) y el enfriamiento por calentamiento son los principales cuellos de botella en la informática). Una CPU está manejando muchas preguntas al mismo tiempo o está siendo arrastrada en muchas direcciones, está descubriendo qué hacer para no poder hacerlo. (oye, es casi humano)

Una gpu es un trabajador gruñón, el trabajo tedioso. Una CPU está gestionando un caos completo y no puede manejar cada detalle.

Entonces, ¿qué aprendemos? Una gpu detalla todo el trabajo tedioso a la vez y una cpu es una máquina de tareas múltiples que no puede enfocarse muy bien con demasiadas tareas que hacer. (Es como si tuviera trastorno de atención y autismo al mismo tiempo).

Ingeniería allí son las ideas, el diseño, la realidad y mucho trabajo duro.

Cuando me vaya, recuerde comenzar de manera simple, comenzar rápidamente, fallar rápidamente, fallar rápido y nunca dejar de intentarlo.

Andrew G. Corbi
fuente