In Thrust we code

Librería Thrust
Clase 5
http://fisica.cab.cnea.gov.ar/gpgpu/index.php/en/icnpg/clases
Thrust
●
●
●
Vectors
Iterators and Static
Dispatching
Algorithms
– Transformations
●
●
–
Reductions
–
Prefix-Sums
●
–
Reordering
●
–
Sorting
–
Etc.
Repaso, y un poco de C++
Fancy Iterators
– constant_iterator
–
transform_iterator
–
permutation_iterator
–
zip_iterator
Best Practices
Thrust 1.8
¿ Que es Thrust ?
http://thrust.github.io/
excelente documentación, muchos ejemplos
Repaso de Thrust
Hasta aqui...
Un poco de C++
Function Templates
Functors
Generic Algorithms
Algoritmos, tipos y operadores
Operadores predefinidos
Ejemplo de “device backends” y reducción
con operador personalizado
●
Tareas:
–
hacer la suma de N numeros random, usando
thrust::generate y thrust::reduce [LISTO].
–
Encontrar el mínimo de N numeros random,
usando thrust::generate y thrust::reduce.
device_backends.cu
[koltona@compute-0-0 intro_thrust]$ more device_backends_compilation.sh
nvcc -Xcompiler -fopenmp device_backends.cu -DTHRUST_DEVICE_SYSTEM=THRUST_DEVICE_SYSTEM_CPP -o
cpp.out
nvcc -Xcompiler -fopenmp device_backends.cu -DTHRUST_DEVICE_SYSTEM=THRUST_DEVICE_SYSTEM_CUDA -o
cuda.out
nvcc -Xcompiler -fopenmp device_backends.cu -DTHRUST_DEVICE_SYSTEM=THRUST_DEVICE_SYSTEM_OMP -o
omp.out
qsub submit_cuda.sh
qsub submit_omp.sh
qsub submit_cpp.sh
Operadores propios
comparadores
Predicados
Fancy Iterators
– constant_iterator
– transform_iterator
– permutation_iterator
– zip_iterator
Quick-Start-Guide THRUST
Fancy Iterators
constant_iterator
Simplemente un iterador que retorna el mismo valor, que uno define, cada vez
que uno lo dereferencia:
#include <thrust/iterator/constant_iterator.h>
...
// create iterators
thrust::constant_iterator<int> first(10);
thrust::constant_iterator<int> last = first + 3;
first[0]
// returns 10
first[1]
// returns 10
first[100] // returns 10
// sum of [first, last)
thrust::reduce(first, last);
// returns 30 (i.e. 3 * 10)
constant_iterator es una solución eficiente y conveniente
cada vez que uno necesita una secuencia de valores
constantes.
Quick-Start-Guide THRUST
Fancy Iterators
transform_iterator
Nos permite hacer “Kernel Fusion”: combinar dos algoritmos en uno solo, como hace
transform_reduce, pero mas customizado:
#include <thrust/iterator/transform_iterator.h>
// initialize vector
thrust::device_vector<int> vec(3);
vec[0] = 10; vec[1] = 20; vec[2] = 30;
// create iterator (type omitted)
... first = thrust::make_transform_iterator(vec.begin(), negate<int>());
... last = thrust::make_transform_iterator(vec.end(),
negate<int>());
first[0]
// returns -10
first[1]
// returns -20
first[2]
// returns -30
// sum of [first, last)
thrust::reduce(first, last);
// returns -60 (i.e. -10 + -20 + -30)
Se puede omitir la declaracion de los iteradores:
// sum of [first, last)
thrust::reduce(thrust::make_transform_iterator(vec.begin(), negate<int>()),
thrust::make_transform_iterator(vec.end(),
negate<int>()));
Quick-Start-Guide THRUST
Fancy Iterators
permutation_iterator
#include <thrust/iterator/permutation_iterator.h>
...
// gather locations
thrust::device_vector<int> map(4);
Map[0] = 3; map[1] = 1; map[2] = 0; map[3] = 5;
// array to gather from
thrust::device_vector<int> source(6);
source[0] = 10; source[1] = 20; source[2] = 30; source[3] = 40;
source[4] = 50; source[5] = 60;
// fuse gather with reduction:
//
sum = source[map[0]] + source[map[1]] + ...
int sum =
thrust::reduce(thrust::make_permutation_iterator(source.begin(),
map.begin()),thrust::make_permutation_iterator(source.begin(),
map.end()));
Nos permite fusionar scatter y gather
Quick-Start-Guide THRUST
Aplicacion del permutation iterator
Imagen “encriptada”
Imagen “desencriptada”
Los pixels estan mezclados usando un mapeo dado...
[FINAL ICNPG2014]
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Kernel Fusion
[koltona@gpgpu-fisica ejemplos]$ cp
/share/apps/codigos/thrust_examples/xxxxx.cu .
[koltona@gpgpu-fisica ejemplos]$ nvcc xxxxx.cu
[koltona@gpgpu-fisica ejemplos]$ qsub script.sh
[koltona@gpgpu-fisica ejemplos]$ more
script.sh
#! /bin/bash
#
#$ -cwd
#$ -j y
#$ -S /bin/bash
## pido la cola gpu.q
#$ -q gpu.q
## pido una placa
#$ -l gpu=1
#
#ejecuto el binario
./a.out
hacerenclase.cu
Con transform_reduce, o transform_iterator
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Ejemplos de la librería
[koltona@gpgpu-fisica ejemplos]$ cp
/share/apps/codigos/thrust_examples/xxxxx.cu .
[koltona@gpgpu-fisica ejemplos]$ nvcc xxxxx.cu
[koltona@gpgpu-fisica ejemplos]$ qsub script.sh
[koltona@gpgpu-fisica ejemplos]$ more
script.sh
#! /bin/bash
#
#$ -cwd
#$ -j y
#$ -S /bin/bash
## pido la cola gpu.q
#$ -q gpu.q
## pido una placa
#$ -l gpu=1
#
#ejecuto el binario
./a.out
sorting_aos_vs_soa.cu
Thrust Best Practices
Thrust by Example, GTC, Julien Demouth, Nvidia
Ejemplos de la librería
[koltona@gpgpu-fisica ejemplos]$ cp
/share/apps/codigos/thrust_examples/xxxxx.cu .
[koltona@gpgpu-fisica ejemplos]$ nvcc xxxxx.cu
[koltona@gpgpu-fisica ejemplos]$ qsub script.sh
[koltona@gpgpu-fisica ejemplos]$ more
script.sh
#! /bin/bash
#
#$ -cwd
#$ -j y
#$ -S /bin/bash
## pido la cola gpu.q
#$ -q gpu.q
## pido una placa
#$ -l gpu=1
#
#ejecuto el binario
./a.out
counting_iterator.cu
permutation_iterator.cu
Hibrid programming
Using a system-specific vector
Execution policies
Ejemplo de execution policies
alumnos_icnpg2015/intro_thrust/clase2/omp_vs_cuda/ompvsgpu_same_host.cu
●
nvcc -Xcompiler -fopenmp ompvsgpu_same_host.cu
●
qsub submit.sh
●
Ver output...
Using a system-specific vector
Paralelismo de datos y de tareas
CPU esperando...
Paralelismo de datos y de tareas
farber1-11_sync.cu
alumnos_icnpg2015/intro_thrust/clase2/omp_vs_cuda/ompvsgpu_same_host.cu
Paralelismo de datos y de tareas
farber1-11_sync.cu
farber1-11_async.cu
PROBAR!
Thrust 1.8 (el ultimo)
Thrust v1.8.0 release
We are pleased to announce the release of Thrust v1.8, an open-source C++
library for developing high-performance parallel applications. Modeled after the
C++ Standard Template Library, Thrust brings a familiar abstraction layer to the
realm of parallel computing.
Thrust 1.8.0 (viene incluido con CUDA 7):
support for algorithm invocation from CUDA __device__ code
support for CUDA streams
Algorithm performance improvements
thrust::complex template provides complex number support .
Volveremos con Thrust despues de unas clases...

Download Report