1. No category

Aceleración en GPU explotando
paralelismo de datos
Indice de contenidos [50 diapositivas]
Curso de Verano de la UMA “Programando la GPU con CUDA”
Málaga, del 15 al 24 de Julio de 2015
1. Introducción a la aceleración en GPU. [5]
2. Alternativas para la programación de GPUs. [37]
1.
2.
3.
4.
Librerías y herramientas para CUDA. [10]
Orientar CUDA hacia otras plataformas. [5]
Cómo acceder a CUDA desde otros lenguajes. [6]
Uso de directivas: OpenACC. [11]
3. Ejemplos de programación: Seis formas de implementar
SAXPY en GPU. [9]
Manuel Ujaldón
Profesor Titular @ Universidad de Málaga
Conjoint Senior Lecturer @ Univ. de Newcastle (Australia)
CUDA Fellow @ Nvidia
2
242 aplicaciones muy populares
aceleradas en GPU
I. Introducción a la aceleración en GPU
www.nvidia.com/appscatalog
4
El mejor rendimiento en aplicaciones científicas
Escalabilidad:
Las aplicaciones escalan a miles de GPUs
Aceleración de una GPU Tesla K20X respecto a una CPU Sandy Bridge
Ingeniería: MATLAB (FFT)*
Física: Chroma
Geología: SPECFEM3D
Dinámica molecular: AMBER
Escalabilidad débil
Reultados en CPU: Dual socket E5-2687w, 3.10 GHz. Resultados en GPU: Dual socket E5-2687w + 2 GPUs Tesla K20X
*Los resultados en MATLAB comparan una CPU i7-2600K frente a una GPU Tesla K20
Advertencia:Las implementaciones que no corresponden a Nvidia pueden no estar completamente optimizadas
Escalabilidad fuerte
5
6
Consumo:
Otro aspecto clave en supercomputación
2.3 PFLOPS
7000 hogares
7.0
Megavatios
7.0
Megavatios
Las CPUs tradicionales no
son económicamente viables
CPU
Optimizada
para tareas
secuenciales
GPU
Optimizada
para multitud
de tareas paralelas
10x performance/socket
> 5x energy efficiency
II. Alternativas de programación
La era de la computación
acelerada en GPUs está aquí
7
CUDA se dirige tanto a
educación como a investigación...
... y proporciona una amplia gama de
herramientas para la rápida adopción de GPUs
Paquetes+
numéricos
MATLAB
MathemaAca
NI+LabView
pyCUDA
Depuradores
&+Profilers
cuda0gdb
Visual+Profiler+
Parallel+Nsight+
Visual+Studio
Allinea
TotalView
Compiladores
C
C++
Fortran
OpenCL
DirectCompute
Java
Python
Paralelizadores
PGI+Accelerator
CAPS+HMPP
mCUDA
OpenMP
BLAS
FFT
LAPACK
NPP
Sparse
Imaging
RNG
Soluciones+OEM
Consultoría+y+entrenamiento
ANEO
Librerías
GPU Tech
9
CUDA soporta un ecosistema muy diverso
en torno a la programación paralela
10
Plataforma de computación paralela CUDA
Este ecosistema puede
articularse en 4 vertientes
principales:
Librerías)
Aceleración
subordinada
La cadena de herramientas
de compilación.
Los lenguajes de
programación.
Las librerías de código.
Las herramientas de
desarrollo.
Direc5vas)
OpenACC)
Lenguajes)de)
programación)
Aceleración fácil
de aplicaciones
Máxima flexibilidad
Nsight IDE
Linux, Mac y Windows
Depuración y
profiling en GPU
Depurador
CUDA-GDB
NVIDIA Visual
Profiler
Habilita la compilación de otros lenguajes a la
plataforma CUDA, y los lenguajes CUDA a otro HW.
SMX
11
Paralelismo
dinámico
Hyper-Q
GPUDirect
12
Entorno de desarrollo Nvidia Nsight,
edición Visual Studio
Anteriormente conocida como Nvidia Parallel Nsight.
Herramienta para la depuración y optimización
(debugging & profiling) integrada en MS Visual Studio.
Disponible para Windows Vista, Windows 7 y Windows 8.
Permite depuración remota y local para:
CUDA C/C++, Direct Compute, HLSL.
Localizador de errores en acceso a memoria
(memchecker)
CUDA C/C++.
II.1. Entornos, librerías y
herramientas para CUDA
Optimización (profiling) para:
CUDA C/C++, OpenCL, Direct Compute.
14
Depuración con Nvidia Nsight
Profiling con Nvidia Nsight
15
15
16
Librerías: Sencillez y aceleración de calidad
Tres pasos para acelerar aplicaciones CUDA
Facilidad de uso: Utilizar librerías habilita la aceleración
en GPU sin tener conocimientos avanzados de programación.
"Drop-in": Muchas librerías aceleradas en GPU siguen
APIs estándar, permitiendo acelerar con mínimos cambios.
Calidad: Las librerías ofrecen implementaciones de alta
calidad para funciones que aparecen en un amplio rango de
aplicaciones.
Rendimiento: Las librerías de Nvidia han sido
optimizadas por expertos.
Paso 1: Sustituir las llamadas a librerías por las llamadas a
librerías CUDA equivalentes.
saxpy(...) -->
cublasSaxpy (...)
Paso 2: Gestionar la localidad de datos.
Con CUDA:
cudaMalloc(), cudaMemcpy(), etc.
Con CUBLAS: cublasAlloc(), cublasSetVector(), etc.
Paso 3: Reconstruir y enlazar la librería acelerada en
CUDA.
nvcc myobj.o -l cublas
17
Un ejemplo de álgebra lineal: saxpy.
for(i=0;i<N;i++) y[i]=a*x[i]+y[i];
int N = 1 << 20;
cublasInit();
cublasAlloc(N, sizeof(float), (void**)&d_x);
cublasAlloc(N, sizeof(float), (void**)&d_y);
cublasSetVector(N, sizeof(x[0]), x, 1, d_x, 1);
cublasSetVector(N, sizeof(x[0]), y, 1, d_y, 1);
// Computa saxpy sobre un millón de elementos
cublasSaxpy(N, 2.0, d_x, d_y, 1);
Librerías matemáticas para CUDA
Rutinas matemáticas de altas prestaciones para sus
aplicaciones:
Inicializa CUBLAS
Aloja vectores en la GPU
cuFFT: Librerías con transformadas rápidas de Fourier.
cuBLAS: Librería BLAS (Basic Linear Algebra Subroutines) íntegra.
cuSPARSE: Librería para matrices dispersas.
cuRAND: Librería para la generación de números aleatorios.
NPP: Primitivas de rendimiento para procesar imágenes y vídeo.
Thrust: Plantillas para algoritmos paralelos y estructuras de datos.
math.h: Librería C99 de punto flotante.
Transfiere datos a GPU
Coloca el prefijo "cublas"
y utiliza las vars. de GPU
cublasGetVector(N, sizeof(y[0]), d_y, 1, y, 1);
Devuelve los datos
desde la GPU
cublasFree(d_x);
cublasFree(d_y);
Desaloja vectores
en la GPU
cublasShutdown();
Desactiva la
librería CUBLAS
18
Incluidas todas ellas en el toolkit de CUDA, descarga
gratuita en: https://developer.nvidia.com/cuda-downloads
19
20
Librerías y herramientas: El ecosistema del
programador permite crecer sus aplicaciones
Librerías aceleradas en GPU
Otras muchas otras librerías fuera del CUDA toolkit...
Desarrolladas por Nvidia.
De código abierto.
NVIDIA cuBLAS
NVIDIA cuRAND
Vector Signal
Image Processing
GPU Accelerated
Linear Algebra
Matrix Algebra
on GPU and
Multicore
IMSL Library
Building-block
ArrayFire Matrix
Algorithms
for
Computations
CUDA
Sparse Linear
Algebra
NVIDIA cuSPARSE
NVIDIA NPP
NVIDIA cuFFT
C++ STL
Features for
CUDA
... además de la ingente cantidad de programas disponibles en la red gracias al esfuerzo de tantos programadores.
Descritas en detalle en Nvidia Developer Zone:
21
http://developer.nvidia.com/cuda-tools-ecosystem
22
Compilando para otras plataformas
II. 2. Cómo orientar CUDA
hacia otras plataformas
24
Ocelot
http://code.google.com/p/gpuocelot
Swan (última versión: Diciembre de 2010)
http://www.multiscalelab.org/swan
Es un entorno de compilación
dinámico para el código PTX
sobre sistemas heterogéneos
que permite un análisis
extensivo del mismo y
su migración a otras
plataformas.
Es un traductor fuente-a-fuente de CUDA a OpenCL:
Proporciona una API común que abstrae el soporte en tiempo de
ejecución de CUDA y OpenCL.
Preserva la comodidad de lanzar kernels CUDA
(<<<bloques,hilos>>>), generando código fuente C para el punto de
entrada a las funciones kernel.
... pero el proceso de conversión precisa de intervención humana.
Útil para:
Evaluar el rendimiento de un código OpenCL para un código CUDA.
Reducir la dependencia de nvcc cuando compilamos código host.
Desde Feb'11 (versión 2.0),
también considera:
GPUs de AMD/ATI.
CPUs x86 de Intel.
25
MCUDA
http://impact.crhc.illinois.edu/mcuda.php
Proporcionar múltiples CUDA Compute Capability en un solo binario.
Como librería run-time para gestionar kernels OpenCL en nuevos
desarrollos.
26
El compilador CUDA x86 del Portland Group
www.pgroup.com/resources/cuda-x86.htm
Desarrollado por el grupo de investigación IMPACT en la
Universidad de Illinois.
Es un entorno de trabajo basado en Linux que trata de
migrar códigos CUDA eficientemente sobre CPUs multicore.
Disponible para su descarga gratuita ...
Principales diferencias con las herramientas anteriores:
En lugar de traducir desde código fuente, trabaja en tiempo de
ejecución. Esto permite construir código binario unificado que
simplifica la distribución software.
Principales ventajas:
Velocidad: La versión compilada puede ejecutarse en una
plataforma x86 incluso sin disponer de GPU. El compilador vectoriza
el código para las instrucciones SSE (128 bits) y AVX (256 bits).
Transparencia: Incluso aquellas aplicaciones que utilizan recursos
nativos como las unidades de textura tendrán un comportamiento
idéntico tanto en CPU como en GPU.
Disponibilidad: Licencia gratis un mes registrándose como CUDA
developer.
27
28
Lenguajes de programación más populares
II. 3. Cómo acceder a CUDA
desde otros lenguajes
Fuente: CodeEval.com [2013]. Muestra superior a 100.000 códigos.
30
CUDA & Python:
Una alianza de rendimiento y productividad
Hay un punto de entrada a CUDA desde los
lenguajes de programación más populares
Python es el lenguage de mayor crecimiento. ¿Por qué?
Los seis perfiles de programador más usuales disponen
hoy en día de herramientas para conectar con CUDA.
Resulta fácil de aprender.
Cuenta con potentes librerías.
Ofrece una simbiosis entre empresas y programadores: Las firmas
plantean desafíos y los desarrolladores envían sus resultados.
Python marca también la tendencia en la contratación:
Buena parte de los programadores Python están buscando trabajo.
PyCUDA permite acceder al API de CUDA desde Python:
Mapea la totalidad de CUDA sobre Python. Documentación completa.
Generación de código en run-time (RTCG) para una rápida y
automática optimización de códigos.
Robustez (errores), comodidad (librerías), agilidad de los wrappers.
31
1. Programador de C
2. Programador de Fortran
CUDA C, OpenACC.
CUDA Fortran, OpenACC.
3. Programador de C++
4. Programador matemático
Thrust, CUDA C++.
MATLAB, Mathematica, LabVIEW.
5. Programador de C#
6. Programador de Python
GPU.NET.
PyCUDA.
32
Envolventes y generadores de interfaz
a través de llamadas C/C++
La creciente diversidad de lenguajes de
programación con los que interactúa CUDA
CUDA puede incorporarse dentro de cualquier lenguaje que
admita llamadas a rutinas C/C++. Para simplificar el proceso,
podemos utilizar generadores de interfaz de propósito general.
SWIG [http://swig.org] (Simplified Wrapper and Interface
Generator) es la aproximación más popular a este respecto.
Está activamente soportada e incluye una larga lista de
lenguajes (ver próxima diapositiva).
También se dispone de una conexión con interfaz Matlab:
Para una GPU: Utilizar Jacket (plataforma de computación numérica).
Para múltiples GPUs: Utilizar MatWorks Parallel Computing Toolbox.
33
Lenguajes orientados a dominios concretos
(DSL = Domain-Specific Languages)
34
Puedes comenzar hoy mismo
Muy populares dentro de su área de aplicación específica:
Matemáticas, estadística, física, medicina, ...
Todos esos lenguajes son soportados por las GPUs CUDA.
Seguramente ya tienes una GPU de ésas en tu portátil o
sobremesa (como pieza aparte o integrada con la CPU).
Páginas Web:
CUDA C/C++: http://developer.nvidia.com/cuda-toolkit
CUDA Fortran: http://developer.nvidia.com/cuda-toolkit
Thrust C++ Template Lib: http://developer.nvidia.com/thrust
GPU.NET: http://tidepowerd.com
PyCUDA (Python): http://developer.nvidia.com/pycuda
MATLAB: http://www.mathworks.com/discovery/matlab-gpu.html
Mathematica: http://www.wolfram.com/mathematica/new-in-8/
cuda-and-opencl-support
35
36
Código fuente del compilador disponible:
Una puerta abierta a lenguajes y desarrollos
Conexión sencilla y
eficiente con un sinfín de:
Lenguajes por encima, por
ejemplo, agregando front-ends
para Java, Python, R, DSLs.
Arquitecturas por debajo,
por ejemplo, ARM, FPGA, x86.
Nvidia modificó LLVM
para construir sobre él un
nvcc, posibilitando:
CUDA
C, C++, Fortran
Cobertura de
nuevos lenguajes
Compilador LLVM para CUDA
GPUs de
Nvidia
CPUs x86
Publicar el código fuente.
Enriquecerse de las aportaciones de la comunidad LLVM.
Cobertura de
nuevo hardware
II. 4. Uso de directivas: OpenACC
37
OpenACC: Un esfuerzo corporativo
para la estandarización
OpenACC: Una alternativa a la computación
CUDA para programadores menos avezados
Se trata de un estándar para la programación paralela de
aceleradores basada en directivas (tipo OpenMP), que:
Se insertan en programas C, C++ o Fortran.
Orientan al compilador para paralelizar ciertas secciones de código.
Objetivo: Pensadas para un programador no experto.
El código es portable a procesadores multicore y paralelos.
Primeros desarrollos e iniciativas comerciales:
The Portland Group (PGI).
Cray.
Primeros clientes en el campo de la supercomputación:
39
Estados Unidos: Oak Ridge National Lab.
Europa: Swiss National Supercomputing Centre.
40
OpenACC: Directivas
OpenACC: Cómo funcionan las directivas
Las directivas proporcionan la base para un código
común que es:
Partiendo de sencillas
anotaciones, el compilador
paraleliza el código.
Cuantas más anotaciones
coloques, mejor será la
generación de código del
compilador.
Funciona en:
Multi-plataforma.
Multi-fabricante.
Esto ofrece una forma abierta de preservar la inversión
en aplicaciones antiguas, habilitando una fácil migración
hacia la computación basada en aceleradores hardware.
Las directivas GPU permiten beneficiarnos de toda la
potencia de la GPU de una manera sencilla, sobre todo si
lo comparamos con los hilos de CPU o los kernels de CUDA.
El mayor logro es no tener que reestructurar el
código cuando queramos paralelizar una gran aplicación.
Los dos pasos básicos para comenzar
Paso 1: Anotar las directivas en el código fuente.
!$acc data copy(util1,util2,util3) copyin(ip,scp2,scp2i)
!$acc parallel loop
… <código fuente>
!$acc end parallel
!$acc end data
Paso 2: Compilar y ejecutar.
pgf90 -ta=nvidia -Minfo=accel file.f
GPUs many-core.
CPUs multi-core.
CPU
GPU
Program myscience
... serial code ...
!$acc kernels
do k = 1,n1
do i = 1,n2
... parallel code ...
enddo
enddo
!$acc end kernels
...
End Program myscience
Anotación
para el
compilador
OpenACC
Código C o
Fortran original
41
42
Un ejemplo
!$acc data copy(A,Anew)
iter=0
do while ( err > tol .and. iter < iter_max )
Copiar los vectores en la
memoria de la GPU dentro
de la región de datos
iter = iter +1
err=0._fp_kind
!$acc kernels
do j=1,m
do i=1,n
Anew(i,j) = .25_fp_kind *( A(i+1,j ) + A(i-1,j ) &
+A(i ,j-1) + A(i ,j+1))
err = max( err, Anew(i,j)-A(i,j))
end do
end do
!$acc end kernels
Paralelizar el código
interno a la región
Delimitar la región paralela
IF (mod(iter,100)==0 .or. iter == 1) print *, iter, err
A= Anew
end do
!$acc end data
Delimitar la región de datos
y devolver sus valores
44
La pregunta clave es:
¿Cuánto rendimiento sacrificamos?
Ejemplos más recientes
Algunos resultados indican pérdidas de sólo el 5-10%
frente a CUDA en "algunos" casos concretos. Otras fuentes
ganan un factor 5x dedicando una semana o incluso un día.
En realidad, este factor depende más de la bondad de la
aplicación que de la habilidad del programador.
Detección de objetos en
tiempo real
Valoración de acciones con
métodos de MonteCarlo
Interacción entre solventes
y biomoléculas
Global Manufacturer of Navigation Systems
Global Technology Consulting Company
University of Texas at San Antonio
El ciclo de vida
del pescado
en Australia
University of Melbourne
65x en 2 días
5x en 40 horas
2x en 4 horas
5x en 8 horas
Estrellas y galaxias Redes neuronales
hace 12.5B años
en auto-aprendiUniversity of Groningen
zaje de robots
The University of Plymouth
5.6x en 5 días
4.7x en 4 horas
45
Un testimonio de un reciente seminario de
OpenACC en Pittsburgh Supercomp. Center
46
Más casos estudios del congreso GTC'13:
3 compiladores OpenACC [PGI, Cray, CAPS]
Rendimiento en la GPU M2050 (Fermi, 14x 32 cores),
sin contar el tiempo de transferencia entre CPU y GPU.
Tamaño de las matrices a multiplicar: 2048x2048.
Stencil de 7 puntos. Tamaño del vector 3D: 256x256x256.
Al+final+del+segundo+día
10x+en+un+kernel+atmosférico
incorporando+sólo+6+direcAvas
Technology+Director
NaAonal+Center+for+Atmospheric+
Research+(NCAR)
47
Fuente: "CUDA vs. OpenACC:
Performance Case Studies",
por T. Hoshino, N. Maruyama,
S. Matsuoka.
48
Para comenzar a programar con OpenACC
Regístrate para obtener acceso libre al compilador de
directivas por un mes, e incluso herramientas para una
rápida adaptación (ver http://www.nvidia.com/gpudirectives)
También hay un compilador de CAPS por 199 €.
III. Ejemplos de programación:
Seis formas de implementar SAXPY en GPU
49
¿Qué es SAXPY? Single-precision Alpha X
Plus Y. Forma parte de la librería BLAS.
1. CUDA C
Código C estándar
void saxpy_serial(int n, float a, float *x, float *y)
{
for (int i = 0; i < n; ++i)
y[i] = a*x[i] + y[i];
}
// Invocar al kernel SAXPY secuencial (1M elementos)
saxpy_serial(4096*256, 2.0, x, y);
Utilizando este código básico, ilustraremos seis formas
diferentes de programar la GPU:
Código CUDA equivalente de ejecución paralela en GPU:
CUDA C.
CUBLAS Library.
CUDA Fortran.
Thrust C++ Template Library.
C# con GPU.NET.
OpenACC.
__global__ void saxpy_parallel(int n,float a,float *x,float *y)
{
int i = blockIdx.x*blockDim.x + threadIdx.x;
if (i < n) y[i] = a*x[i] + y[i];
}
// Invocar al kernel SAXPY paralelo (4096 bloques de 256 hilos)
saxpy_parallel<<<4096, 256>>>(4096*256, 2.0, x, y);
51
52
2. CUBLAS Library
3. CUDA Fortran
Fortran estándar
Código BLAS secuencial
int N = 1 << 20;
// Utiliza la librería BLAS de tu elección
// Invoca a la rutina SAXPY secuencial (1M elementos)
blas_saxpy(4096*256, 2.0, x, 1, y, 1);
Código cuBLAS paralelo
int N = 1 << 20;
cublasInit();
cublasSetVector (N, sizeof(x[0]), x, 1, d_x, 1);
cublasSetVector (N, sizeof(y[0]), y, 1, d_y, 1);
// Invoca a la rutina SAXPY paralela (1M elementos)
cublasSaxpy (N, 2.0, d_x, 1, d_y, 1);
cublasGetVector (N, sizeof(y[0], d_y, 1, y, 1);
cublasShutdown();
Fortran paralelo
module my module contains
subroutine saxpy (n, a, x, y)
real :: x(:), y(:), a
integer :: n, i
do i=1,n
y(i) = a*x(i) + y(i);
enddo
end subroutine saxpy
end module mymodule
module mymodule contains
attributes(global) subroutine saxpy(n, a, x, y)
real :: x(:), y(:), a
integer :: n, i
attributes(value) :: a, n
i = threadIdx%x + (blockIdx%x-1) * blockDim%x
if (i<=n) y(i) = a*x(i) + y(i)
end subroutine saxpy
end module mymodule
program main
use mymodule
real :: x(2**20), y(2**20)
x = 1.0, y = 2.0
program main
use cudafor; use mymodule
real, device :: x_d(2**20), y_d(2**20)
x_d = 1.0, y_d = 2.0
$ Invoca SAXPY para 1M elementos
call saxpy(2**20, 2.0, x, y)
end program main
$ Invoca SAXPY para 1M elementos
call saxpy<<<4096,256>>>(2**20, 2.0, x_d, y_d)
y = y_d
end program main
53
4.1. CUDA C++:
Desarrollar código paralelo genérico
54
4.2. Thrust C++ STL
Con CUDA C++ podemos desarrollar código paralelo
genérico, que permite encarar aplicaciones sofisticadas y
flexibles con un rico middleware:
Thrust es una librería de algoritmos paralelos y código
abierto que reestructura la librería STL (Standard Template
Library) de C++. Principales características:
Jerarquía de clases.
Métodos __device__.
Plantillas (templates).
Sobrecarga de operadores.
Functors (objetos función).
New/delete en device.
...
Interfaz de alto nivel:
Mejora la productividad del programador.
Habilita la portabilidad del rendimiento entre GPUs y CPUs.
Flexible:
Para back-ends de CUDA, OpenMP y Thread Building Blocks (TBB).
Extensible y ajustable a medida.
Se integra con el software existente.
Eficiente:
Código GPU escrito sin emplear directamente llamadas a kernels CUDA.
55
56
4.2. Thrust C++ STL (cont.)
C# estándar
Código C++ secuencial
con STL y Boost
int N = 1<<20;
std::vector<float> x(N), y(N);
...
5. C# con GPU.NET
Código C++ paralelo
private static
void saxpy (int n, float a,
float[] a, float[] y)
int N = 1<<20;
thrust::host_vector<float> x(N), y(N);
...
{
for (int i=0; i<n; i++)
y[i] = a*x[i] + y[i];
...
thrust::device_vector<float> d_x = x;
thrust::device_vector<float> d_y = y;
}
// Invocar SAXPY para 1M elementos
// Invocar SAXPY para 1M elementos
std::transform(x.begin(), x.end(),
thrust::transform(x.begin(), x.end(),
y.begin(), x.end(),
y.begin(), y.begin(),
2.0f * _1 + _2);
2.0f * _1 + _2);
int N = 1<<20;
// Invoca SAXPY para 1M elementos
saxpy(N, 2.0, x, y)
C# paralelo
[kernel]
private static
void saxpy (int n, float a,
float[] a, float[] y)
{
int i = BlockIndex.x * BlockDimension.x +
ThreadIndex.x;
if (i < n)
y[i] = a*x[i] + y[i];
}
int N = 1<<20;
Launcher.SetGridSize(4096);
Launcher.SetBlockSize(256);
// Invoca SAXPY para 1M elementos
saxpy(2**20, 2.0, x, y)
http://www.boost.org/libs/lambda
http://developer.nvidia.com/thrust
57
6. Directivas de compilación OpenACC
Código C paralelo
void saxpy (int n, float a,
float[] a, float[] y)
{
#pragma acc kernels
for (int i=0; i<n; i++)
y[i] = a*x[i] + y[i];
}
...
// Invoca SAXPY para 1M elementos
saxpy(1<<20, 2.0, x, y)
...
58
Síntesis final
Código Fortran Paralelo
Estas 6 formas de programar están soportadas en todas
las GPUs que aceptan CUDA (más de 600 millones en 2015).
Seguramente tienes una en tu PC portátil o sobremesa.
subroutine saxpy(n, a, x, y)
real :: x(:), y(:), a
integer :: n, i
$!acc kernels
do i=1. n
y(i) = a*x(i) + y(i)
enddo
$!acc end kernels
end subroutine saxpy
1. CUDA C/C++
2. CUDA Fortran
http://developer.nvidia.com/cuda-toolkit
http://developer.nvidia.com/cuda-fortran
...
3. CUBLAS Library
4. Thrust
$ Invoca SAXPY para 1M elementos
call saxpy(2**20, 2.0, x_d, y_d)
...
http://developer.nvidia.com/cublas
http://developer.nvidia.com/thrust
5. C# con GPU.NET
6. OpenACC
http://tidepowerd.com
http://developer.nvidia.com/openacc
59
60