Processamento de Alto Desempenho utilizando Unidade de ... · PCI Bus Host Device. PCI Bus Host Device 1. Copiardados da memóriada CPU para a memóriada GPU 2. ... int *c) {c[

Francisco Ribacionka e Ettore Enrico (STI ) – USP/STI/InterNuvem – [email protected]

Março - 2016

Processamento de Alto Desempenho utilizando

Unidade de Processamento Gráfico - GPU

Programação

1. Utilização de GPU para HPC

• O que é processamento por GPU

• Recursos para desenvolvimento de software

• Aplicações

2. Tesla K20m

3. HPC com recursos da InterNuvem: Cluster lince

4. Computação heterogênea

• Blocos

• Threads



Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



Aplicação

+

Código Paralelo

Código Sequencial

Cache

ULAControle

ULA

ULA

ULA

DRAM

DRAM

GPU CPU

Código Sequencial

1. Copiar dados da memória da CPU

para a memória da GPU

PCI Bus

HostDevice

PCI Bus

HostDevice



2. Carregar o código GPU e executar

PCI Bus

HostDevice



2. Carregar o código GPU e executar

3. Copiar os resultados da memória da

GPU para a memória da CPU

Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



NVIDIA - CUDA

• CUDA é uma plataforma de computação paralela e um modelo de

programação inventados pela NVIDIA

• Linguagens de Programação C, C++, Python, Fortran

CUDA-MPI

• Cuda-MPI: Suporte a chamadas MPI a partir da memória da GPU

OpenCL

• OpenCL: Suporte a Open Computing Language através do CUDA

OpenACC

• OpenACC : Acelerando as aplicações através de diretivas de programação

Fonte: https://nvidia.developer.com

Bibliotecas Otimizadas para GPU’s

• Thrust (C++ Template Library);

• cuBLAS (Algebra linear);

• cuSPARSE (Algebra linear);

• NPP (Processamento de imagem e sinais);

• cuFFT (Processamento de imagem e sinais).

Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



Aplicações:

Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



Tesla K20m

GPU 1xGK110

Precisão Simples 3.5 TF

Precisão Dupla 1.2 TF

Memória 4.8 GB

Largura de Banda 208 GB/s

# Cores CUDA 2496

Energia 225W

Cluster Jaguar – SGI Altix 450

• 56 CPUs Intel Itanium 2

• 112 GB de RAM

• 0,34 TFlops

Cluster Puma

• Servidores DeLL - Rede Gibabit-ethernet

•59 X 16 GB de memória RAM

•59 X 16 CPUs Intel Xeon 2.6 GHz

• 1.26 TFlops

Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



•32 servidores, cada um com:

• 2 Tesla k20m

• 16 cores Intel E7-2870 (2,4 GHz)

• 129 GB RAM

•Scratch com 55 TB

•Sistema de filas Torque

Computação de alto desempenho – HPC

Softwares Instalados na lince que utilizam GPU:

• gromacs

• lammps

• NAMD

Sugestões: [email protected]

#!/bin/bash -v

#PBS -S /bin/bash

#PBS -N cuda

#PBS -l nodes=1:ppn=16

#PBS -l gpus=2

#PBS -joe

#PBS -l walltime= 249:00:00

#PBS -q parallel

ulimit -s unlimited

module load cuda/7.0

cd /scratch/fribacio/1_Utilities/deviceQuery

time ./deviceQuery

Job CUDA:

Recursos

Fila

Comandos para

executar o

programa

$ qsub job_cuda.sh

7927.lince.lcca.usp.br

$ qstat

lince.lcca.usp.br:

Req'd Req'd Elap

Job ID Username Queue Jobname SessID NDS TSK Memory Time S Time

-------------- -------------- -------- ------------- ------ ----- ------ ----------- ---------- - ------------

7927.lince fribacio parallel cuda 795 1 16 -- 3000:00:0 R 00:00:00

$ ls

cuda.o7927

CUDA Device Query (Runtime API) version (CUDART static linking)

Detected 2 CUDA Capable device(s)

Device 0: "Tesla K20m"

CUDA Driver Version / Runtime Version 7.0 / 6.5

CUDA Capability Major/Minor version number: 3.5

Total amount of global memory: 4800 MBytes (5032706048 bytes)

(13) Multiprocessors, (192) CUDA Cores/MP: 2496 CUDA Cores

GPU Clock rate: 706 MHz (0.71 GHz)

Memory Clock rate: 2600 Mhz

Memory Bus Width: 320-bit

L2 Cache Size: 1310720 bytes

.... ...

Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



Computação Heterogênea

� Terminologia:

� Host CPU e a memória RAM

� Device GPU e a sua memória

Host Device

24

NVIDIA – CUDA

• CUDA é uma plataforma de computação paralela e um modelo

de programação inventados pela NVIDIA

• Linguagens de Programação C, C++, Python, Fortran

• Baseado no padrão C

• Extensões que permitem programação heterogênea

• APIs para gerenciar memória e dispositivos

Fonte: https://nvidia.developer.com

Computação Heterogênea

__global__ void mykernel(void) {

}

int main(void) {

mykernel<<<1,1>>>();

printf(“STI-USP\n");

return 0;

}$ nvcc sti_usp.cu

$ a.out

STI-USP

$

26

Soma de dois inteiros:

#include <stdio.h>

__global__ void add( int a, int b, int *c ) { *c = a + b;}

int main( void ) {

int c; int *dev_c;

cudaMalloc( (void**)&dev_c, sizeof(int) );

add<<<1,1>>>( 2, 7, dev_c );

cudaMemcpy( &c, dev_c, sizeof(int),

cudaMemcpyDeviceToHost );

printf( "2 + 7 = %d\n", c );

cudaFree( dev_c ); return 0;

}27

Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



Programação Paralela em CUDA

int main( void ) {

int a[N], b[N], c[N];

for (int i=0; i<N; i++) {

a[i] = -i;

b[i] = i * i;

}

add( a, b, c );


printf( "%d + %d = %d\n", a[i], b[i], c[i] );

}

return 0;

}

#include <stdio.h>

#define N 10

void add( int *a, int *b, int *c ) {

int indice = 0;

while (indice < N) {

c[indice] = a[indice] + b[indice];

indice += 1;

}

}

Versão Host

30

#include <stdio.h>

#define N 10

__global__ void add( int *a, int *b, int *c ) {

int indice = blockIdx.x;

if (indice < N)

c[indice] = a[indice] + b[indice];

}

int main( void ) {

int a[N], b[N], c[N];

int *dev_a, *dev_b, *dev_c;

cudaMalloc( (void**)&dev_a, N * sizeof(int) ) ;

cudaMalloc( (void**)&dev_b, N * sizeof(int) ) ;

cudaMalloc( (void**)&dev_c, N * sizeof(int) ) ;


a[i] = -i;

b[i] = i * i;

}

cudaMemcpy( dev_a, a, N * sizeof(int),

cudaMemcpyHostToDevice ) ;

cudaMemcpy( dev_b, b, N * sizeof(int),

cudaMemcpyHostToDevice ) ;

add<<<N,1>>>( dev_a, dev_b, dev_c );

cudaMemcpy( c, dev_c, N * sizeof(int),

cudaMemcpyDeviceToHost ) ;


printf( "%d + %d = %d\n", a[i], b[i], c[i] );

}

cudaFree( dev_a ) ;

cudaFree( dev_b ) ;

cudaFree( dev_c ) ;

return 0;

}

Versão Device

31

• Computação com GPU é para paralelismo massivo– Como rodar código em paralelo no device?

add<<< 1, 1 >>>();

add<<< N, 1 >>>();

• Em vez de executar add() uma vez, executar N vezes emparalelo


32

__global__ void add(int *a, int *b, int *c) {

c[blockIdx.x] = a[blockIdx.x] + b[blockIdx.x];

}

• Na GPU, cada bloco pode executar em paralelo:

c[0] = a[0] + b[0]; c[1] = a[1] + b[1]; c[2] = a[2] + b[2]; c[3] = a[3] + b[3];

Block 0 Block 1 Block 2 Block 3


33

Blocos e Grids

Programação




• Aplicações

2. Tesla K20m



• Blocos

• Threads



int tid = threadIdx.x + blockIdx.x * blockDim.x;

add<<<4,4>>>( dev_a, dev_b, dev_c );

int tid = threadIdx.x + blockIdx.x * blockDim.x;

Hierarquia de Memória

Hierarquia de thread

• threadIdx é um vetor de 3 componentes

• Formando blocos de threads de uma, duas ou três

dimensões

• Provendo uma maneira natural para computar vetores,

matrizes ou volumes

__global__ void MatAdd(float A[N][N], float B[N][N], float C[N][N])

{ int i = threadIdx.x;

int j = threadIdx.y;

C[i][j] = A[i][j] + B[i][j];

}

Int main()

{

// Chamada kernel de um bloco com N * N * 1 threads

int numBlocks = 1;

dim3 threadsPerBlock(N, N);

matAdd<<<numBlocks, threadsPerblock>>> (A, B, C);

...

Modelo de Hardware

SMX: 192

single-precision

CUDA cores, 64

double-precision

units, 32 special

function units (SFU),

and 32 load/store

units (LD/ST).

Tesla K20m

Multiprocessor

count: 13

Shared mem per

mp: 49152

Registers per

mp: 65536

Bibliografia

• SANDERS, J.; KANDROT, E. CUDA by Example - An Introduction to

General-Purpose GPU Programming. Addison-Wesley. 2011

• KIRK, D.; HWU, W. Programming Massively Parallel Processors - A

Hands-on Approach. Morgan Kaufmann. 2010



Dúvidas? [email protected]

Francisco Ribacionka e Ettore Enrico (STI ) – USP/STI/InterNuvem – [email protected]

Março - 2016



Documents

Processamento de Alto Desempenho utilizando Unidade de ... · PCI Bus Host Device. PCI Bus Host Device 1. Copiardados da memóriada CPU para a memóriada GPU 2. ... int *c) {c[