Programación heterogénea y oneAPI Toolkit. Conferencia improvisada experta de Intel responde sus preguntas



Como parte de la columna "Haga una pregunta a un experto de Intel", le pedimos al especialista líder de Intel, Konstantin Vladimirov, que respondiera preguntas relacionadas con la programación heterogénea, el kit de herramientas oneAPI y cosas interesantes relacionadas. El resultado superó todas nuestras expectativas. Konstantin no perdió tiempo y dio respuestas detalladas y fundamentadas, sin temor a ser polémico. De hecho, obtuvimos una pequeña conferencia sobre programación de arquitectura cruzada en todas sus formas: matices de descarga, optimizaciones, estándares, etc.
Transferimos el micrófono al experto. Bueno, los comentarios se dan a la audiencia.

imagen Pregunta Soarex16
¿Qué tan laboriosa será la transición de OpenCL a oneAPI y qué beneficios se pueden obtener de esto?

La respuesta Cambiar a DPC ++ puede ser complicado, pero en mi opinión, vale la pena. Hay dos etapas principales.

En primer lugar, esta es una transición desde su lenguaje de programación heterogéneo (OpenCL, Vulkan compute), que, muy probablemente, se basa en la API. Aquí tiene una ventaja en el hecho de que ya conoce el área temática, y la dificultad está en cambiar el pensamiento del control directo a través de la API a construcciones de lenguaje ligeramente más implícitas.
En segundo lugar, esta es una transición desde su idioma de host. Si ha estado descargando toda su vida de C puro, entonces el umbral de entrada es igual al umbral para cambiar de C a C ++, que es bastante alto.

¿Por qué intentarlo?

En primer lugar, DPC ++ hace un gran trabajo para un programador. Muy rápidamente olvidará, como una pesadilla, todas estas llamadas explícitas a clXXXYYY, y lo que significa el sexto argumento, y si olvidó el código de retorno. Muchos contenedores orientados a objetos ocultan la rutina no peor, pero generalmente a costa de cambiar de la API estándar de OpenCL a la API de contenedor no tan estándar (también vi esas bicicletas). En el caso de DPC ++, simplemente escriba el SYCL estándar con extensiones Intel (que pronto también se convertirá en el SYCL estándar).

En segundo lugar, DPC ++ proporciona una compilación conjunta, es decir, puede estar seguro de los tipos y no tendrá problemas en los bordes de la API con dimensiones, relleno, alineación. Escribe el núcleo y el código de host en un archivo, y este es el mismo código. Con USM, también puede trabajar con estructuras de datos complejas mucho más fácil.

En tercer lugar, DPC ++ es C ++ real, es decir, permite una programación generalizada. Por ejemplo, el núcleo más simple para agregar dos vectores:

auto kern = [A, B, C](cl::sycl::id<1> wiID) { C[wiID] = A[wiID] + B[wiID]; //   A, B  C?  ! }; 

Lo mismo en OpenCL:

 _kernel void vector_add(__global int *A, __global int *B, __global int *C) { int i = get_global_id(0); C[i] = A[i] + B[i]; } 

Verá, me vi obligado a señalar un tipo OpenCL int. Si necesito un flotante, tendré que escribir otro kernel o usar un preprocesador o generación de código externo. Tener casi todas las características de C ++ a su disposición puede ser un poco aterrador si no ha tenido experiencia con C ++. Pero esto es algo común cuando se trata de un cambio tecnológico importante.

Y todos los beneficios no se limitan a esto. Mencionaré algo más en las siguientes respuestas.

Así que habría descargado el compilador en su lugar y lo había probado, ya que no es difícil hacer esto con el paquete OneAPI .

imagen Pregunta Juster
¿OpenVINO y oneAPI estarán relacionados de alguna manera?

La respuesta La distribución OpenVINO ahora es parte de la distribución OneAPI. Aprender y usar redes neuronales es una tarea computacionalmente difícil que se beneficia enormemente de la programación heterogénea. Creo que tarde o temprano todos los componentes de OneAPI permitirán utilizar todos los recursos informáticos disponibles: aceleradores gráficos y aceleradores especiales como Nervana y FPGA. Y todo esto sin abandonar el paradigma del lenguaje y el sistema de tipos de su programa C ++.

imagen Preguntas por correo
Estoy tratando de entender cómo se verá el acelerador de hardware de IA en 3 años, por favor ayuda con esto. Hay una compañía interesante Graphcore y su IPU: este dispositivo no es menos eficiente que FPGA, pero es mucho más fácil de programar: Python con soporte para TensorFlow y otros marcos. Resulta que si se cumplen las promesas de Graphcore, no habrá necesidad de FPGA en el mercado de aprendizaje automático. Python es mucho más conveniente para los científicos de datos que C ++.
¿Acepta que FPGA no es adecuado para el mercado de aprendizaje automático en comparación con las soluciones programables de Python? Si se pierde el mercado de ML, ¿qué otras aplicaciones FPGA generalizadas ves?
¿En qué aplicaciones ve la necesidad inevitable de una programación heterogénea, en la que no puede pasar con herramientas más convenientes como Python?

La respuesta Miré brevemente qué tipo de UIP. Una pieza más de hierro en la que todos descargarán. Estos muchachos compiten con la GPU y con aceleradores especiales, y no con la FPGA.

En tareas para las que se afila una pieza de hardware especializada, siempre vencerá a FPGA, por ejemplo, renderizar video es mejor en una tarjeta de video, etc. Pero en el mundo (incluso en el mundo de ML) hay muchas tareas para las cuales no se ha inventado ni lanzado nada especial, y aquí FPGA siempre será indispensable. Por ejemplo, porque hay una cuestión de precio y, para ser barato, una pieza de hardware especializada debe ser masiva.

Supongamos ahora que la IPU especificada es realmente genial. Esto no cancelará la programación heterogénea, por el contrario, la presencia de un acelerador tan excelente lo estimulará. Y también le dará una ventaja enorme a OneAPI y DPC ++, porque tarde o temprano alguien dirá "Quiero usar tanto su IPU como mi GPU desde un programa". Más bien temprano porque la programación heterogénea se trata de eso. Su significado es la descarga de una tarea adecuada a un dispositivo adecuado. Una tarea puede venir de cualquier parte. Y este dispositivo puede ser cualquier cosa, incluso puede ser el mismo dispositivo en el que se ejecuta el programa. Por ejemplo, si descarga el núcleo escrito en ISPC y utiliza las características vectoriales de Xeon al máximo, puede descargarlo usted mismo y seguir siendo una ganancia significativa. El criterio principal aquí es el rendimiento. Bueno, nunca habrá demasiada productividad en este mundo. Incluso con los mejores aceleradores del mundo.

En cuanto a Python y su conveniencia ... Tengo que admitir de inmediato que no me gustan los lenguajes escritos dinámicamente: son lentos y, en lugar de un error de compilación normal, debe esperar dos horas antes de caer en el tiempo de ejecución debido al tipo incorrecto. Pero no veo lo malo que es hacer las mismas descargas de Python. Por cierto, OneAPI ya incluye Intel Distribution para Python, lo cual es extremadamente conveniente para varias revisiones.

Es decir, en el mundo de los sueños de los amantes de Python, usted escribe un programa en él y lo descarga a todos los aceleradores que puede encontrar usando OneAPI, y no a un montón de bibliotecas específicas del proveedor. Otra cosa es que con este enfoque, se pierde la escritura de extremo a extremo y se regresa al mundo extremadamente precario de la programación basada en API. Quizás el desarrollo de DPC ++ alentará a la comunidad a usar más activamente herramientas más apropiadas, como C ++.

imagen Pregunta del correo
Rendimiento versus OpenCL. Debe haber impuestos sobre el lujo, es decir gastos generales ¿Hay alguna medida?

La respuesta En Internet puede encontrar muchas mediciones con una variedad de resultados, dependiendo del compilador, la tarea y la calidad de la implementación. Como investigación personal, medí tareas simples (SGEMM, DGEMM) en mi computadora portátil (gráficos integrados de Skylake), y vi que hasta ahora hay una reducción (dentro del porcentaje). Pero me parece que esto es una consecuencia del hecho de que todo esto es beta hasta ahora.

En teoría, el resultado debería ser la aceleración, no la desaceleración, es decir, en principio, todo este lujo debería tener un valor negativo. Se trata del compilador. Cuando su programa consta de una sola fuente y se procesa como un solo programa, el compilador obtiene oportunidades fantásticas e increíbles para la optimización: diseño de código común, inversión de bucles, reorganización de secciones de código y todo lo demás que el compilador simplemente no puede hacer en el enfoque basado en API, pero tarde o temprano, definitivamente aprenderá con un modelo de fuente única.

Además, DPC ++ tendrá un costo negativo en términos de tiempo de desarrollo. Un ejemplo simple son los accesos SYCL, que el compilador ya está utilizando para organizar eventos y administrar colas asíncronas.

 deviceQueue.submit([&](cl::sycl::handler &cgh) { auto A = bufferA.template get_access<sycl_read>(cgh); auto B = bufferB.template get_access<sycl_read>(cgh); auto C = bufferC.template get_access<sycl_write>(cgh); .... deviceQueue.submit([&](cl::sycl::handler &cgh) { auto A = bufferA.template get_access<sycl_read>(cgh); auto B = bufferB.template get_access<sycl_read>(cgh); auto D = bufferD.template get_access<sycl_write>(cgh); 

Aquí, el compilador ve que ambos paquetes solo leen A y B y escriben buffers independientes C y D, como resultado, él ve la capacidad de enviarlos en paralelo si hay suficientes tamaños globales.

Por supuesto, un programa OpenCL escrito pedagógicamente puede hacerlo igual de bien, pero el tiempo de desarrollo empleado con un núcleo no trivial no será comparable.

imagen Pregunta del correo
¿Son relevantes todas las formas de optimizar las aplicaciones OpenCL para DPC ++? ¿Qué novedades hay que agregarles?

La respuesta Diría que la mayor parte de la sutil optimización manual que realizan los escritores de kernel puede y debe realizarla el compilador. De la misma manera, por ejemplo, considero una práctica perjudicial instalar manualmente un ensamblador en línea en programas C ++, porque incluso si proporciona beneficios tácticos, interfiere con las optimizaciones y actúa como un factor negativo en el desarrollo y la transferencia de un producto. Bueno, OpenCL ahora también es ensamblador.

En cuanto a la respuesta más detallada, tengo miedo del abismo aquí. Por ejemplo, hay un conocido documento de Intel "OpenCL Developer Guide for Intel Processor Graphics". Y hay una sección sobre cómo intentarlo, para no poner donde está el exceso de sincronización.

Entonces, desde mi punto de vista, esta es una tarea no humana en principio. Las personas son extremadamente pobres en razonamiento sobre la sincronización de subprocesos múltiples y tienden a esculpir la sincronización de forma conservadora o incorrecta, o ambas a la vez: pongo comas como esta ( pero lo arreglamos: nota editorial ).

Por otro lado, en DPC ++, en lugar de escribir código con barreras explícitas, así:

  for (t = 0; t < numTiles; t++) { const int tiledRow = TS * t + row; const int tiledCol = TS * t + col; Asub[col][row] = A[globalRow * AY + tiledCol]; Bsub[col][row] = B[tiledRow * BY + globalCol]; // Synchronise to make sure the tile is loaded barrier(CLK_LOCAL_MEM_FENCE); // .... etc .... 

Lo más probable es que escriba una iteración explícita de parallel_for_work_group , dentro de la cual group.parallel_for_work_item

 cgh.parallel_for_work_group<class mxm_kernel>( cl::sycl::range<2>{BIG_AX / TS, BIG_BY / TS}, cl::sycl::range<2>{TS, TS}, [=](cl::sycl::group<2> group) { // .... etc .... for (int t = 0; t < numTiles; t++) { group.parallel_for_work_item([&](cl::sycl::h_item<2> it) { // .... etc .... Asub[col][row] = A[globalRow][tiledCol]; Bsub[col][row] = B[tiledRow][globalCol]; }); //      ,    

Como resultado, no tiene que configurar la sincronización a mano en absoluto, y se puede descartar toda la sección.

Y así puedes caminar en todas las secciones. Algo sobrevivirá, algo se irá. Preveo la aparición de un nuevo documento "Optimización para DPC ++", pero el tiempo debería pasar, ya que todas las técnicas que funcionan realmente se desarrollan solo más tarde y con sangre

imagen Pregunta del correo
Hay una limitación en OpenCL: no puede usar "datos distantes" en el núcleo, es decir, implementar un "filtro ancho" que use datos de entrada de un gran grupo de píxeles más grande que el grupo de trabajo OpenCL en un cálculo. ¿Qué ofrece DPC ++ a este respecto?

La respuesta Bueno, eso es imposible. Por supuesto, no escribo núcleos en particular ... Pero es absolutamente seguro que puede usar toda la memoria global tal como está, solo necesita asegurarse de trabajar con operaciones atómicas (o sincronizar núcleos jerárquicos externamente). Y también puede conectar System SVM (bueno, o hay USM en DPC ++).

Por desgracia, todo esto es extremadamente ineficiente, y no me gustan todos estos trucos. Además, son difíciles de optimizar por el compilador.

Y así, si hablamos de soluciones directas y efectivas, entonces, por supuesto, no hay magia en DPC ++. Al final, su programa todavía se divide en partes: el código de host y el código del dispositivo, y todas las restricciones del dispositivo afectan el código del dispositivo. El tamaño máximo del grupo de trabajo es ese paralelismo real que su hardware es capaz de hacer. Todo lo que está encima de esto son solo formas de salir, afectando dramáticamente negativamente el rendimiento. Es por eso que DPC ++ brinda la oportunidad de hacer esto: device.get_info <sycl :: info :: device :: max_work_group_size> () y luego decidir cómo vivir con el número resultante.

Sería tentador, por supuesto, hacer un modelo en DPC ++, cuando el programador trabaja como desee con bucles de cualquier longitud, y el compilador analiza qué hacer a continuación, pero sería mortal, porque ocultaría constantes e incluso asintóticos de complejidad adicional. informática que aparece de la nada. Por otra razón, Alexandrescu escribió que "encapsular la complejidad debería considerarse un delito", y esto también se aplica.

A veces es útil revisar el algoritmo en sí. Aquí DPC ++ facilita las cosas porque un código más estructurado es más fácil de refactorizar. Pero esto es más o menos un consuelo.

imagen Pregunta del correo
DPC ++ se basa en SYCL. Pero qué pasa si profundiza bajo el capó, cuáles son las diferencias con OpenCL en la implementación del back-end, si corresponde. Por ejemplo, ¿el mecanismo de distribución entre dispositivos heterogéneos es el mismo que OpenCL?

La respuesta Si te pones bajo el capó, entonces esto es OpenCL. Todas las ventajas y fortalezas de SYCL son las ventajas y fortalezas del lenguaje, es decir, la interfaz. Desde el extremo frontal viene el viejo SPIRV que va al back-end y allí está optimizado (a menudo ya en tiempo de ejecución, es decir, es JIT) ya para una tarjeta de video específica de la misma manera que OpenCL estaría optimizado para ello.

Otra cosa es que el mecanismo para distribuir el trabajo entre dispositivos heterogéneos es más front-end que back-end, porque es el código de host el que decide qué enviar y dónde. Y el código de host se obtiene de DPC ++. Ya he mostrado un ejemplo un poco más alto de cómo el compilador puede, en base a los accesores, tomar una decisión sobre la paralelización de paquetes. Y esto es solo la punta del iceberg.

imagen Pregunta del correo
Bibliotecas Sí, no estamos hablando de CUDA. Pero sabemos que para los desarrolladores de CUDA hay bibliotecas muy útiles que funcionan con alto rendimiento en la GPU. OneAPI también contiene algunas bibliotecas, pero, por ejemplo, IPP : no hay ningún archivo útil para trabajar con imágenes en oneAPI / OpenCL. ¿Habrá algo y cómo, en este caso, cambiar de CUDA a oneAPI?

La respuesta La transición de CUDA a un único estándar abierto será difícil, pero inevitable. Por supuesto, CUDA ahora tiene una infraestructura más madura. Pero las características de su licencia son un inconveniente de bloqueo, porque cada vez aparecen más jugadores en el mercado de sistemas heterogéneos, más y más tarjetas interesantes y aceleradores de diferentes fabricantes.

La diversidad de las API existentes dificulta el uso de este mundo de posibilidades para los programadores con experiencia en la CPU clásica. Lo que lleva a OneAPI o algo así. Aquí la magia no está en el avance de Intel en los gráficos, sino en el hecho de que Intel abre la puerta a DPC ++ para todos. Ni siquiera poseemos el estándar SYCL, pertenece al grupo Khronos y todas las extensiones de Intel son extensiones en Khronos donde cualquiera puede comprometerse (y hay representantes de todos los principales jugadores allí). Y eso significa que aparecerán (bibliotecas) y la comunidad (ya están apareciendo), y un montón de vacantes en esta dirección.

Y, por supuesto, IPP se reescribirá para nuevas realidades. No tengo nada que ver con IPP, pero usar DPC ++ es de sentido común, y hay personas sensatas sentadas allí.

Pero lo más importante, ahora es el momento en la historia en el que puede escribir su propia biblioteca, que superará el IPP y que todo el mundo usará. Porque los estándares abiertos siempre ganan.

imagen Pregunta del correo
Si comparamos el lanzamiento de algoritmos de entrenamiento y de inferencia de redes neuronales en Nervana y FPGA, ¿cuáles son las diferencias en la programación y la eficiencia resultante?

La respuesta No sé nada sobre los detalles de programación de FPGA, escribo compiladores. Pero tengo una contrapregunta. ¿Y cómo vamos a comparar? En los puntos de referencia estándar es antideportivo, Nervana lamió debajo de ellos. Pero en caso de que tenga algo interesante, entonces el FPGA le desatará las manos, y poner esto en Nervana puede ser largo y costoso, eso es todo.

Resulta que la pregunta en sí es, por así decirlo, de la serie "quién es más fuerte que un elefante o una ballena". Pero esta no es una pregunta real. La verdadera pregunta es: ¿cómo aprovechar un elefante y una ballena en un solo carro? Bueno, o al menos distribuir, digamos, para que un elefante lo arrastre por tierra y una ballena por mar.

En el caso de OneAPI, tendrá el mismo programa, en general, en C ++ estándar. Y puede escribirlo usted mismo y ejecutarlo con descarga de un lado a otro. Esta será la tarea que le interesa, en la que usted mismo puede medir y optimizar el rendimiento. Un único estándar y una única interfaz para dispositivos heterogéneos será un paso hacia la comparación de manzanas con manzanas en tales asuntos.

Por ejemplo: "¿Qué es mejor para el% de mi tarea% desde el punto de vista de la facilidad de programación y la eficiencia: poner esta parte en FPGA, dejar esta en Nervana o dividir esta parte en dos y reescribir esta parte para la GPU?"

Y toda la historia con OneAPI: es solo para que usted diga: "por qué pensarlo durante mucho tiempo, lo intentaré ahora rápidamente, ES SENCILLO".

Todavía no, no es fácil. Pero lo habrá.



Epílogo del experto

Gracias a todos por sus preguntas. Es posible e incluso probable que me haya equivocado, sea inexacto y haya cometido errores. Sucede, en Internet constantemente alguien está equivocado.

Espero haber podido interesar a alguien en programación heterogénea y DPC ++. Quiero recomendar a todos el sitio sycl.tech , donde se encuentran toneladas de informes, incluso de expertos de renombre mundial (se requiere inglés)

Bueno para todos!

PD del editor. Esta vez, por decisión unánime del consejo editorial, se decidió otorgar el premio a la mejor pregunta ... al autor de las respuestas. Creo que estará de acuerdo en que esto es justo.

Source: https://habr.com/ru/post/479486/


All Articles