GPU Computing vs. CPU Multinuclear: Análise de desempenho para processamento de iluminação por Ray Tracing em modelos tridimensionais

Arthur Pereira Gregorio

Rodrigo Zanatta Rodriguez

GPU Computing vs. CPU Multinuclear: Analise de desempenho para

processamento de iluminacao por Ray Tracing em modelos tridimensionais

FOZ DO IGUACU - PR

2011

CESUFOZ - CENTRO DE ENSINO SUPERIOR DE FOZ DO IGUACU

PARANA





Trabalho de Conclusao de Curso sub-

metido ao Centro de Ensino Superior

de Foz do Iguacu como parte dos re-

quisitos para obtencao do grau de

bacharel em Ciencia da Computacao.

Orientador: Gildomiro Bairros.

FOZ DO IGUACU - PR

2011





Este exemplar corresponde a redacao final do

trabalho de conclusao de curso apresentado

como requisito parcial para obtencao do grau

de Bacharel em Ciencia da Computacao, CE-

SUFOZ - Centro de Ensino Superior de Foz

do Iguacu, aprovada pela comissao formada

pelos professores:

Gildomiro Bairros, MSc.Orientador

Weber Takaki, Esp.Avaliador Cesufoz

Itamar Pena Nieradka, Esp.Avaliador Cesufoz

Foz do Iguacu, 25 de Novembro de 2011

DEDICATORIA

Dedicamos este trabalho as nossas famılias

e amigos, quem nos apoiou e incentivou

quando achamos que tudo estava perdido...

AGRADECIMENTOS

Agradecemos inicialmente ao professor Gildomiro Bairros pelo apoio durante a

orientacao, ao Dr. Thanassis Tsiodras pelo seu esplendido trabalho na criacao da

Renderer utilizada, ao grupo GPU Brasil pelo apoio na solucao de duvidas sobre a

plataforma CUDA, nossas famılias e amigos. O apoio de todos foi indispensavel durante

realizacao deste trabalho.

“Nove pessoas nao podem fa-

zer um bebe em um mes.”

Fred Brooks

Resumo

Com a crescente demanda por processamento de alta performance e paralelismo, foinecessario que criassem formas de aproveitar um dispositivo presente em muitas maquinasatuais e normalmente subutilizado: a GPU.

Neste trabalho sera realizada uma comparacao de desempenho entre as plataformasGPGPU e CPU multinuclear utilizando as interfaces de programacao paralela CUDA eINTEL TBB para renderizacao de objetos tridimensionais iluminados por ray tracing.

Keywords: General-Purpose Computation on Graphics Processing Units, Raytra-cing, CUDA.

Abstract

With the growing demand for high processing performance and parallelism, it wasnecessary to create ways to get a device present in many current machines and typicallyunderutilized: GPU.

In this study, we conducted a performance comparison between platforms using mul-ticore CPU and GPGPU programming interfaces for parallel CUDA and Intel TBB forrendering three-dimensional objects illuminated by ray tracing.

Keywords: General-Purpose Computation on Graphics Processing Units, Raytra-cing, CUDA.

Lista de Abreviaturas e Siglas

ALU Arithmetic Logic Unit

API Application Programming Interface

CU Control Unit

CUDA Compute Unified Device Architecture

FPS Frames Per Second

GB Gigabytes

GCC GNU Compiler Collection

Ghz Giga-hertz

GPGPU General-Purpose computation on Graphics Processing Units

Intel TBB Intel Threading Building Blocks

Mhz Mega-hertz

MIMD Multiple Instruction Multiple Data

MISD Multiple Instruction Single Data

MLAA Morphological Anti Aliasing

OpenCL Open Computing Language

OpenGL Open Graphics Library

OpenMP Open Multi-Processing

SDL Simple DirectMedia Layer

SIMD Single Instruction Multiple Data

SIMT Single Instruction Multiple Threads

SISD Single Instruction Single Data

Lista de Figuras

2.1 Comparacao evolutiva entre CPUs e GPUs na capacidade processamento

de operacoes de ponto flutuante . . . . . . . . . . . . . . . . . . . . . . . . 6

2.2 Fluxo de trabalho do pipeline programavel . . . . . . . . . . . . . . . . . . 8

2.3 CPU vs. GPU: arquitetura interna . . . . . . . . . . . . . . . . . . . . . . 10

2.4 Ray Tracing, como funciona . . . . . . . . . . . . . . . . . . . . . . . . . . 13

2.5 Processamento Serial . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 14

2.6 Processamento Paralelo . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 14

2.7 Segmentos do mercado que mais utilizam computacao paralela . . . . . . . 15

2.8 Estrutura de camadas da arquitetura CUDA . . . . . . . . . . . . . . . . . 16

2.9 Sequencia de execucao do aplicativo entre a devide e o host . . . . . . . . . 19

2.10 Grids de thread blocks do modelo de programacao CUDA . . . . . . . . . . 21

4.11 Modelo de complexidade simples, Chessboard . . . . . . . . . . . . . . . . 30

4.12 Modelo de complexidade media, Dodge Challenger 2007 . . . . . . . . . . . 30

4.13 Modelo de complexidade alta, Happy . . . . . . . . . . . . . . . . . . . . . 31

4.14 Tempo de renderizacao em cada cenario para o modelo de baixa complexidade 33

4.15 Taxa de FPS para o modelo de baixa complexidade . . . . . . . . . . . . . 33

4.16 Tempo de renderizacao em cada cenario para o modelo de media complexidade 34

4.17 Taxa de FPS para o modelo de media complexidade . . . . . . . . . . . . . 35

4.18 Tempo de renderizacao em cada cenario para o modelo de alta complexidade 36

4.19 Taxa de FPS para o modelo de alta complexidade . . . . . . . . . . . . . . 36

Lista de Quadros

2.1 Exemplo da chamada de um kernel . . . . . . . . . . . . . . . . . . . . . . 20

2.2 Codigo de um aplicativo do tipo hello world escrito em CUDA. Fonte:

Galbraith (2009) . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 24

Lista de Tabelas

2.1 Taxonomia de Flynn de acordo com os fluxos de instrucoes e dados . . . . 11

2.2 Tipos de memoria presentes no device . . . . . . . . . . . . . . . . . . . . . 22

2.3 Principais funcoes suportadas para cada versao de compute capability . . . 23

4.4 Resultados obtidos, modelo de baixa complexidade . . . . . . . . . . . . . 32

4.5 Resultados obtidos, modelo de media complexidade . . . . . . . . . . . . . 34

4.6 Resultados obtidos, modelo de alta complexidade . . . . . . . . . . . . . . 35

Sumario

1 Introducao 1

1.1 Objetivos . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 2

1.1.1 Objetivo Geral . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 2

1.1.2 Objetivos Especıficos . . . . . . . . . . . . . . . . . . . . . . . . . . 2

1.2 Metodologia . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 2

2 Fundamentacao Teorica 4

2.1 Central Processing Unit - CPU . . . . . . . . . . . . . . . . . . . . . . . . 4

2.2 Graphics Processing Unit - GPU . . . . . . . . . . . . . . . . . . . . . . . 5

2.2.1 Pipeline Grafico . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 6

2.2.1.1 Pipeline grafico de funcao fixa . . . . . . . . . . . . . . . . 7

2.2.1.2 Pipeline grafico programavel . . . . . . . . . . . . . . . . . 7

2.2.1.3 Shaders Graficos . . . . . . . . . . . . . . . . . . . . . . . 8

2.3 Computacao Heterogenea . . . . . . . . . . . . . . . . . . . . . . . . . . . 9

2.3.1 GPU Computing . . . . . . . . . . . . . . . . . . . . . . . . . . . . 9

2.3.1.1 CPU vs. GPU . . . . . . . . . . . . . . . . . . . . . . . . 9

2.4 Taxonomia de Flynn . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 10

2.5 Renderizacao de Imagens . . . . . . . . . . . . . . . . . . . . . . . . . . . . 11

2.5.1 Renderizacao Tridimensional . . . . . . . . . . . . . . . . . . . . . . 12

2.5.2 Ray Tracing . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 12

2.6 Programacao Paralela . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 13

2.7 Compute Unified Device Architecture - CUDA . . . . . . . . . . . . . . . . 15

2.7.1 Stream Processing . . . . . . . . . . . . . . . . . . . . . . . . . . . 16

2.7.2 C for CUDA . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 16

2.7.3 Modelo de Programacao em CUDA . . . . . . . . . . . . . . . . . . 16

2.7.3.1 A API CUDA . . . . . . . . . . . . . . . . . . . . . . . . . 17

2.7.3.2 Kernels e hierarquia de Threads . . . . . . . . . . . . . . 20

2.7.3.3 Hierarquia de Memoria . . . . . . . . . . . . . . . . . . . . 21

2.7.3.4 CUDA Compute Capability . . . . . . . . . . . . . . . . . 22

2.7.3.5 Codigo de Exemplo . . . . . . . . . . . . . . . . . . . . . . 23

3 Ambiente Experimental 25

3.1 Estrutura Fısica . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 25

3.1.1 Maquina Utilizada . . . . . . . . . . . . . . . . . . . . . . . . . . . 25

3.2 Estrutura Logica . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 26

3.2.1 Sistema Operacional . . . . . . . . . . . . . . . . . . . . . . . . . . 26

3.2.2 Plataforma CUDA . . . . . . . . . . . . . . . . . . . . . . . . . . . 26

3.2.2.1 CUDA Toolkit . . . . . . . . . . . . . . . . . . . . . . . . 26

3.2.3 3Ds MAX 2012 . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 26

3.2.4 Renderer C++ . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 27

3.2.4.1 Simple DirectMedia Layer . . . . . . . . . . . . . . . . . . 27

3.2.5 Renderer CUDA . . . . . . . . . . . . . . . . . . . . . . . . . . . . 27

3.2.5.1 OpenGL . . . . . . . . . . . . . . . . . . . . . . . . . . . . 27

3.2.6 Plataforma C++ . . . . . . . . . . . . . . . . . . . . . . . . . . . . 28

3.2.6.1 Compilador . . . . . . . . . . . . . . . . . . . . . . . . . . 28

3.2.6.2 Intel R© Threading Building Blocks . . . . . . . . . . . . . 28

4 Implementacao 29

4.1 Descricao do Estudo de Caso . . . . . . . . . . . . . . . . . . . . . . . . . . 29

4.1.1 Descricao dos Cenarios de Testes . . . . . . . . . . . . . . . . . . . 31

4.1.1.1 Cenarios . . . . . . . . . . . . . . . . . . . . . . . . . . . . 31

4.2 Resultados . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 32

4.2.1 Modelo de baixa complexidade, Chessboard . . . . . . . . . . . . . 32

4.2.2 Modelo de media complexidade, Dodge Challenger 2007 . . . . . . . 33

4.2.3 Modelo de alta complexidade, Happy . . . . . . . . . . . . . . . . . 35

5 Consideracoes Finais 37

5.1 Problemas Durante os Testes . . . . . . . . . . . . . . . . . . . . . . . . . . 37

5.2 Conclusao . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 38

5.3 Trabalhos Futuros . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . . 38

Referencias 39

1

1 Introducao

Com a evolucao do cenario computacional moderno, sistemas que combinam o uso de

CPU’s de multiplos nucleos e GPU’s de multiplos nucleos trouxeram o processamento em

massa a uma nova escala de desempenho (DAVID B. KIRK, 2010, p.4).

Diversas areas da computacao que antes utilizavam processamento atraves de ser-

vidores compostos por apenas CPU’s multinucleares estao mudando este paradigma e

passando a utilizar o poder dos sistemas heterogeneos a seu favor. Devido a este fato os

sistemas General-Purpose computation on Graphics Processing Units (GPGPU) comecam

a ganhar atencao.

Seguindo a nova tendencia do mercado, em 2007, a nVidia Corporation apresentou

o Compute Unified Device Architecture (CUDA), uma plataforma proprietaria baseada

na linguagem de programacao C/C++ para desenvolvimento paralelo, criada para ser

executada exclusivamente por suas unidades de processamento grafico compatıveis. Logo

o CUDA passou a ser palco de muitas pesquisas relacionadas a mineracao de dados,

pesquisa molecular, calculos complexos como a previsao do tempo, dentre outras.

Um ano depois do lancamento feito pela nVidia, foi apresentado a comunidade open-

source o Open Computing Language (OpenCL), um framework 1 baseado na linguagem de

programacao C para a escrita de aplicativos que executam sobre plataformas heterogeneas,

baseado em paralelismo de dados e tarefas, o OpenCL tornou-se um padrao para este

tipo de desenvolvimento (KHRONOS, 2011a). Mantido pelo consorcio Khronos Group o

OpenCL trouxe uma resposta a arquitetura proprietaria apresentada pela nVidia, com a

proposta de ser livre e compatıvel com qualquer tipo de sistema heterogeneo. Este passa

a ser um dos concorrentes diretos do CUDA.

Com base nesta nesta nova necessidade computacional, este trabalho visa apresentar

uma comparacao de desempenho entre as plataformas compostas por sistemas GPGPU

1Segundo (HORSTMANN, 2006), framework e um conjunto de classes cooperativas que implementamos mecanismos essenciais para um domınio de problemas especıficos.

2

em relacao a um sistema composto apenas por CPU’s multinucleares para o processamento

de calculos de insercao de raios e iluminacao em um cena tridimensional utilizando ray

tracing.

1.1 Objetivos

Nesta secao serao apresentados os objetivos que este trabalho busca alcancar.

1.1.1 Objetivo Geral

Apresentar um comparativo de desempenho que possa indicar a viabilidade do uso de

um sistema GPGPU em relacao a um sistema composto apenas por uma CPU multinuclear

para a rendererizacao de objetos tridimensionais iluminados por ray tracing.

1.1.2 Objetivos Especıficos

• Conceituar sobre programacao paralela e ray tracing ;

• Demonstrar conceitos sobre CPU multinuclear, GPU e plataforma CUDA;

• Apontar as vantagens de se utilizar paralelismo em sistemas GPGPU em relacao a

sistemas com CPU multinucleares;

• Executar medicao de desempenho para o tempo de renderizacao, em segundos, de

objetos tridimensionais utilizando a plataforma GPGPU e CPU multinuclear;

• Executar medicao de desempenho para a taxa Frames Per Second (FPS) durante a

apresentacao das imagens geradas apos a renderizacao dos modelos tridimensionais

nas plataformas GPGPU e CPU multinuclear;

• Apresentar os resultados obtidos tabulados em formato grafico juntamente com os

dados da pesquisa.

1.2 Metodologia

A metodologia aplicada neste trabalho sera com base em revisao bibliografica e analise

quantitativa dos dados gerados pelos testes propostos com base nos cenarios aplicados. Os

3

dados serao apresentados em forma de grafico juntamente com a pesquisa e uma conclusao

sobre o assunto desenvolvido.

4

2 Fundamentacao Teorica

Neste capıtulo serao apresentados os principais conceitos, tecnicas e tecnologias en-

volvidas e necessarias para a compreensao e desenvolvimento deste trabalho.

2.1 Central Processing Unit - CPU

CPU e o hardware responsavel por todo o controle do fluxo de processamento das

informacoes atraves de instrucoes escritas em linguagem de maquina, como por exemplo

o assembly1, onde os programadores podem definir o que ela ira fazer e como ela ira fazer

(BRAIN, 2000).

Com o passar do anos o design das CPU’s mudou drasticamente, hoje existem dispo-

sitivos que oferecem suporte para processamento de 64 bits e multiprocessamento atraves

de multiplos nucleos, sendo eles fısicos ou logicos porem elas ainda continuam sendo com-

postas por duas unidades basicas:

Arithmetic Logic Unit (ALU) local de execucao das operacoes basicas de logica e

matematica

Control Unit (CU) responsavel pela movimentacao dos dados entre a memoria sistema

e a CPU, alem da decodificacao e execucao das informacoes contidas nestes dados.

Seu modo de execucao sequencial baseia-se em quatro etapas: fetch, decode, execute

e writeback sendo que em cada uma delas ocorrem as seguintes operacoes:

• Fetch: perıodo em que a informacao a ser processada e trazida da memoria do

sistema para a memoria da CPU. Este processo e possıvel gracas ao Program Counter

que armazena em um ındice as referencias para o local da memoria dos programas

em execucao (LIN, 2003);

1O mesmo que linguagem de maquina, porem no lugar dos comandos numericos foram colocadossequencias de caracteres para que haja facil memorizacao pelo programador (BARTLETT, 2003, p.5).

5

• Decode: quando atraves do Opcode a CPU define que tipo de operacao sera realizada

com aquele dado especıfico (LIN, 2003);

• Execute: quando a informacao efetivamente esta sendo processada pela CPU. Neste

momento entao a CPU pode invocar o uso da ALU para execucao das operacoes de

logica ou matematica (LIN, 2003);

• Writeback : quando o resultado do processamento e devolvido ao programa em sua

area de memoria (LIN, 2003).

De acordo com Faria (2009), processadores convencionais sao otimizados para cal-

culos sequenciais, fazendo com que estas nao sejam a melhor escolha em um ambiente

massivamente paralelo.

Paralelismo e o futuro da computacao, porem a industria continuara adicionando cores

aos microprocessadores em detrimento ao aumento da performance em single-threads2

(JOHN D. OWENS MIKE HOUSTON, 2008).

2.2 Graphics Processing Unit - GPU

Inicialmente projetadas para processamento grafico, as GPU’s passaram por inumeras

mudancas ate atingir o patamar de desempenho que hoje apresentam (VIANA, 2009).

Com crescimento da industria de jogos as pipelines de funcao fixa passaram a nao

atender mais as necessidades dos programadores. Com o surgimento dos shaders as GPU’s

ganharam a capacidade de processamento sequencial como as CPUs, passando a chamar

atencao dos programadores que buscavam alto poder de paralelismo (FARIA, 2009).

Os pipelines graficos de funcao fixa eram a forma mais utilizada para aplicar transfor-

macoes vetoriais a uma imagem, seja ela tridimensional ou bidimensional, os programado-

res tinham funcoes fixas pre-definidas para uso e nao havia uma forma de personaliza-las.

Normalmente estas funcoes seguiam normas que eram padrao do mercado, tais como

DirectX 3 ou Open Graphics Library (OpenGL)4.

2Tipo de processamento onde apenas um processo e executador por vez em um unico nucleo.3Microsoft DirectX e uma colecao de APIs que tratam de tarefas relacionadas a programacao de

jogos para o sistema operacional Microsoft Windows, e quem padroniza a comunicacao entre software ehardware (HARDWARE CLUBE, 2008)

4Uma biblioteca especificada como padrao para escrita de aplicativos que utilizem graficos em tres ouduas dimensoes, mantido pelo consorcio Khronos Group como uma API Open-Source (KHRONOS, 2011b).

6

Com a chegada dos pipelines programaveis em meados de 2003 o conceito de progra-

macao para GPU’s mudou, pois eles permitiram um maior controle das funcoes graficas

alem da personalizacao sobre a forma como tais funcoes seriam executadas (GUSMaO,

2010).

A figura 2.1 apresenta o crescimento do poder computacional das GPU’s em relacao

as CPU’s entre os anos de 2003 e 2009.

Figura 2.1: Comparacao evolutiva entre CPUs e GPUs na capacidade processamento de ope-racoes de ponto flutuante

Fonte: nVidia (2010b, p.2)

As GPUs possuem realmente um poder muito mais elevado de processamento pois

foram desenvolvidas para executar uma quantidade muito elevada de calculos matematicos

e processamento paralelo intenso (ALVARO CUNO, 2011).

2.2.1 Pipeline Grafico

O pipeline grafico e uma representacao conceitual da forma como os dados sao pa-

ralelamente processados pela GPU. O pipeline grafico, de um modo geral, recebe dados

7

geometricos primitivos (tais como cor, posicao e textura) e os transforma atraves de cal-

culos matematicos para que seja formada a imagem e posteriormente exibida na tela.

2.2.1.1 Pipeline grafico de funcao fixa

O pipeline grafico de funcao fixa ou tambem conhecido como pipeline convencional e

composto por varios estagios, executados atraves de parametros definidos por uma API

previamente associada (IKEDA, 2011).

Estes estagios compreendem:

• Operacoes com vertices: os vertices recebidos pelo pipeline sao convertidos em tri-

angulos;

• Rasterizacao: ato de desenhar linhas de um ponto a outro no espaco ou desenhar

polıgonos a partir de tres ou mais pontos (SANTEE, 2005, p.159);

• Operacoes com Fragmentos: os vertices sao combinados dando origem aos fragmen-

tos, e o passo que mais demanda de poder computacional, pois e nesta etapa em

que a cor final e calculada de forma paralela;

• Composicao: a criacao da imagem final atraves dos fragmentos.

Com funcoes fixas dentro do pipeline sem a possibilidade de personalizacao ou reescrita

durante a programacao, os pipelines de funcao fixa passaram a ser um entrave durante a

elaboracao de efeitos mais complexos.

2.2.1.2 Pipeline grafico programavel

Visando transpor os problemas das funcoes fixas no pipeline de funcao fixa, a industria

trouxe aos programadores os pipelines programaveis.

Com funcoes fixas definidas de fabrica, a criacao de efeitos graficos mais complexos

acabou sendo comprometida. Visando maior flexibilidade e melhor aproveitamento do

hardware grafico, os estagios de manipulacao de vertices e fragmentos passaram a ser to-

talmente programaveis. Houve tabem a adicao de mais um estagio ao pipeline programavel

o geometry shader (IKEDA, 2011).

A figura 2.2 apresenta os estagios de funcionamento de um pipeline programavel.

8

Figura 2.2: Fluxo de trabalho do pipeline programavel

Fonte: adaptado de Wikipedia (2011c)

• Vertex Shader : adiciona os efeitos aos objetos tridimensionais atraves da manipu-

lacao de suas primitivas;

• Geometry Shader : renderiza uma cena e gera novas primitivas a partir dos dados

vindos do processamento realizado pelo vertex shader.

• Pixel Shader : ou Fragment Shader, adiciona os efeitos de cores e iluminacao aos

objetos;

2.2.1.3 Shaders Graficos

Shaders sao pequenos programas que executam dentro de GPU’s ou processadores

compatıveis. Marcaram a transicao entre o uso de pipelines de funcao fixa e o uso de

pipelines programaveis (ZIBULA, 2009).

9

2.3 Computacao Heterogenea

Segundo Sipahi (2009), quando se fala em usar placas de vıdeo e processadores em

conjunto, ou em processadores com varios nucleos que cumprem funcoes diferentes para

processar algum tipo de informacao, estamos falando de Computacao Heterogenea.

De acordo com Sipahi (2009), este conceito da nome a sistemas computacionais onde o

processador ou nucleo central e auxiliado por outros elementos do sistema para a execucao

de um tarefa.

A computacao heterogenea apresenta uma serie de novas fronteiras as serem explora-

das, mas tambem tras problemas que nao existiam quando usada a computacao convenci-

onal. Um destes problemas e a otimizacao de codigo, visto que o software esta anos atras

do hardware no quesito paralelismo. Assim um processador com quatro nucleos pode ter

desempenho inferior a outro de dois nucleos se o software usado nao for otimizado para

tal (FARIA, 2009).

2.3.1 GPU Computing

Segundo Ikeda (2011) e possıvel que alguns autores utilizem o termo GPGPU para

designar o uso de GPU’s em programas de proposito geral, entretanto, o termo mais usual

e o GPU Computing, caracterizando as GPUs totalmente programaveis da ultima geracao.

De acordo com Souza (2011), GPU Computing e a tecnica de uso da GPU para com-

putacao de proposito geral atraves de linguagens ou Application Programming Interface

(API) especıficas para alcancar um alto nıvel de paralelismo em determinados aplicativos.

2.3.1.1 CPU vs. GPU

A constante busca por graficos cada vez mais realistas e processamento em tempo real,

as GPU’s evoluıram para um patamar de processamento onde existem varios nucleos para

execucao de aplicativos altamente paralelos. As CPU’s, devido a sua limitacao fısica e

consumo de energia, acabaram concentrando seu ciclo evolutivo na adicao de mais nucleos

de processamento e nao na melhoria de um em especıfico (IKEDA, 2011).

A figura 2.3 apresenta as diferencas arquiteturais de uma GPU para uma CPU ex-

pondo a maior quantidade de unidades de processamento existentes na unidade grafica,

possibilitando um aumento consideravel no poder computacional. Porem, com numero

reduzido de unidades de controle e cache de dados, o fluxo de execucao do programa acaba

10

se tornado limitado.

Figura 2.3: CPU vs. GPU: arquitetura interna

Fonte: adaptado de nVidia (2010b, p.3)

Ikeda (2011) destaca algumas outras diferencas arquiteturais entre uma CPU e uma

GPU:

• Objetivo: CPU’s tem seu nucleo especializado em executar uma unica thread de

funcoes sequenciais em sua velocidade maxima, ja as GPU’s sao otimizadas para

execucao de milhares de threads paralelamente;

• Multithreading : GPU’s podem manter ate 1024 threads nucleo de processamento,

enquanto as CPU’s podem executar no maximo de 1 a 2 threads por nucleo.

2.4 Taxonomia de Flynn

Segundo Tanenbaum (2006) a taxonomia de computadores paralelos existente ainda

e uma aproximacao muito grosseira da realidade.

A tabela 2.1 apresenta a taxonomia de Flynn, uma das mais utilizadas e coerentes

existentes. Nela a classificacao e feita baseando-se em dois conceitos:

Fluxos de Instrucoes : correspondente ao contador de programas em um sistema que

contenha n CPU’s com n contadores de programas e conseguintemente, n fluxos de

instrucao;

Fluxos de Dados : correspondente a um conjunto de operandos. Por exemplo, um

sensor de temperatura que possui n sensores.

11

Fluxos de Instrucoes Fluxos de Dados Nome Exemplos

1 1 SISD Maquina classica de Von Neumann

1 multiplas SIMD Supercomputador Vetorial

multiplas 1 MISD Nenhum existente

multiplas multiplas MIMD Multiprocessador

Tabela 2.1: Taxonomia de Flynn de acordo com os fluxos de instrucoes e dados

Fonte: adaptado de Tanenbaum (2006, p.345)

Tanenbaum (2006) explica que para cada tipo de denominacao criada por Flynn, existe

um explicacao logica com base em seus fluxos de dados e instrucoes:

• Single Instruction Single Data (SISD): comparavel ao classico computador sequen-

cial de Von Neumann, capaz de executar uma unica operacao por vez obtendo um

unico fluxo de dados por vez;

• Single Instruction Multiple Data (SIMD): possuem uma unica unidade de controle

que emite um fluxo de instrucoes por vez mas que pode ser dividido em varias ALU’s

gerando multiplos fluxos de dados;

• Multiple Instruction Single Data (MISD): nao esta claro que tais maquinas existem,

mas sua taxonomia implica em varias instrucoes operando sobre um unico dado;

• Multiple Instruction Multiple Data (MIMD): multiplas CPU’s operando de forma in-

dependente sobre multiplos conjuntos de dados, mas que fazem parte de um sistema

maior. Um exemplo seriam os clusters.

De acordo com Ikeda (2011) as GPU’s enquadram-se na categoria dos dispositivos

SIMD, pois partindo de uma unica instrucao pode-se obter um numero significativo de

dados.

2.5 Renderizacao de Imagens

Conhecido processo na area de processamento de imagens, a renderizacao compreende

uma tecnica em que modelos graficos tridimensionais ou bidimensionais sao convertidos

em imagens. Neste processo complexo sao realizados inumeros calculos sobre os objetos

presentes na cena, tais como: textura, cor, reflexao, transparencia, posicionamento dos

pontos de iluminacao, sombras e luzes (SILVA BASTOS SALES, 2008).

12

Partindo de arquivos que contenham informacoes sobre as primitivas da cena/imagem

a ser gerada, o processo de renderizacao inicia lendo os dados de geometria, ponto de

visao, texturas, iluminacao e shading5 para entao realizar o processamento, gerando uma

imagem ou o raster grafico6 da imagem/cena.

O processo de renderizacao tambem pode ser aplicado a objetos tridimensionais, os

quais requerem uma maior demanda de processamento.

2.5.1 Renderizacao Tridimensional

Renderizacao tridimensional pode ser definido como o processo de geracao de imagens

atraves de dados tridimensionais dentro de um computador (BIRN, 2002)

A renderizacao tridimensional e um processo de criacao que como tirar um fotografia,

ao final produz algum tipo e imagem sendo que neste a diferenca e que a imagem produzida

possui efeito de visao tridimensional.

O processo de rederizacao tridimensional e composto por basicamente dois estagios:

Reflection/Scattering e Shading. Em cada estagio sao realizadas as seguintes operacoes:

• Reflection/Scattering : definicao de como a luz ira interagir com o objeto partindo

de um ponto em especıfico;

• Shading : definicao de como irao variar as propriedades de iluminacao de um deter-

minado material em relacao a superfıcie e luz que recebe.

2.5.2 Ray Tracing

Ray Tracing e um metodo de renderizacao de iluminacao global, onde e feito o tracado

dos raios de iluminacao para tras do plano da imagem ate os objetos da cena (OWEN, 1999).

Apos a realizacao do processo de tracado dos raios, os raios sao testados em todos

os objetos da cena para determinar se cruzam todos os objetos, se o raio nao passar por

nenhum dos objetos entao o pixel e sombreado pela cor de fundo da imagem.

Na figura 2.4 pode-se ver com detalhes como este processo ocorre, desde a passagem

do raio para o fundo da imagem, chegando ao objeto da cena ate a posicao da camera em

5Processo aplicado durante o desenho de um objeto definindo as areas mais claras e mais escuras emrelacao ao ponto de iluminacao.

6A representacao em mapeamento de bits de uma imagem geralmente mostrada por um grid de pixels(WIKIPEDIA, 2011d).

13

relacao a luz.

Figura 2.4: Ray Tracing, como funciona

Fonte: Wikipedia (2011e)

Silva (2011) define que um algoritmo basico de iluminacao por Ray Tracing e composto

pelas seguintes etapas:

• Calcular uma linha reta unindo o olho do observador ao ponto de visao;

• Descobrir as intersecoes desta reta com os objetos tridimensionais que estao atras

da tela;

• Pintar o ponto com a cor do objeto mais proximo.

Um algoritmo de ray tracing gasta em media 75% do tempo de processamento cal-

culando os pontos de insercao dos raios por envolver calculos matematicos complexos

(WHITTED, 1980).

Devido ao uso intenso de operacoes de ponto flutuante, o processamento de iluminacao

por ray tracing em objetos tridimensionais tem chamado a atencao de pesquisadores da

area de processamento paralelo como GPU Computing.

2.6 Programacao Paralela

Barney (2011) explica que tradicionalmente o software e escrito para execucao serial

e para ser executado em um unico computador com uma unica CPU.

14

Na figura 2.5 pode-se ver um problema computacional sendo divido em pequenas

instrucoes que uma-a-uma sao processadas serialmente.

Figura 2.5: Processamento Serial

Fonte: Barney (2011)

Com o advento dos computadores modernos dotados de CPU’s multinucleares, foi

necessaria a criacao de metodos e formas para que os problemas passem a ser visto de

forma paralela e nao apenas sequencialmente.

Barney (2011) descreve computacao paralela como sendo uma tecnica onde se faz o

uso simultaneo de recursos computacionais para a resolucao de um problema. A figura

2.6 apresenta um cenario computacional paralelo, com varios problemas e varias CPU’s

para que se possa entao realizar o processamento destes de forma paralelizada.

Figura 2.6: Processamento Paralelo

Fonte: Barney (2011)

A computacao paralela segue o paradigma de que o universo e paralelo, onde varios

eventos ocorrem ao mesmo tempo sem que haja uma sequencia entre eles.

A figura 2.7 representa a participacao do processamento paralelo feito atraves de

supercomputadores em cada segmento de mercado durante o ano de 2011. Areas como

15

industria e pesquisa sao as que mais investem nestes sistemas.

Figura 2.7: Segmentos do mercado que mais utilizam computacao paralela

Fonte: Top500 (2011)

2.7 Compute Unified Device Architecture - CUDA

Tomando como base o poder evolutivo das GPU’s, em 2007 a nVidia Corpotation

trouxe ao mercado computacional uma nova arquitetura para programacao atraves de

shaders: o CUDA.

Trazendo as GPU’s o modelo de programacao paralela de proposito geral, o CUDA

iniciou um nova arquitetura de programacao atraves de stream processors disponıveis nas

placas de vıdeo da serie GeForce 8 (IKEDA, 2011).

Abstraindo a complexidade da arquitetura do dispositivo e baseando-se na linguagem

de programacao C como interface de programacao, o CUDA traz aos programadores

uma visao de que a placa e um array de microprocessadores massivamente paralelos

(RITA BORGO, 2009).

Rita Borgo (2009) explica que sua arquitetura de software e composta por varias

camadas, em termos de API, driver de hardware e bibliotecas de utilidades como pode ser

visto na figura 2.8.

16

Figura 2.8: Estrutura de camadas da arquitetura CUDA

Fonte: adaptado de AbiChahla (2008)

2.7.1 Stream Processing

Stream Processing e o modelo de programacao no qual se baseia o CUDA. Tambem

conhecido como Thread Processing, este modelo como finalidade o processamento de um

conjunto uniforme de dados que podem ser operados em paralelo atraves da divisao do

processamento em kernels (IKEDA, 2011).

2.7.2 C for CUDA

A fim de tornar a programacao para GPU algo mais amigavel, o modelo de programa-

cao da API e baseada em uma extensao da linguagem de programacao C (RITA BORGO,

2009).

2.7.3 Modelo de Programacao em CUDA

A estrutura de um programa em CUDA normalmente consiste em duas partes dis-

tintas: codigo a ser executado na CPU e codigo a ser executado na GPU (SCHETINGER,

2009). Neste cenario de programacao, a CPU passa a ser chamada de host ou processador

principal e a GPU device ou coprocessador.

17

Utilizando a CPU como um controlador do fluxo de execucao, a GPU realiza as ope-

racao pesadas sobre os dados utilizando um modelo chamado Single Instruction Multiple

Threads (SIMT) para realizar o processamento de forma totalmente escalar.

O programa sempre sera inicializado a partir da CPU e em algum momento ira agregar

os dados em blocos e passa-los como parametro para uma funcao da GPU (SCHETINGER,

2009).

Schetinger (2009) define que o fluxo basico de execucao de um programa compreende

as seguintes etapas:

• Host inicializa um array com dados;

• Array e copiado da memoria do host para a memoria da device;

• Device opera sobre o vetor de dados;

• O array e copiado novamente para o host.

2.7.3.1 A API CUDA

Para controlar os modelos de threads e memoria, um conjunto de comandos e variaveis

de sistema foi criado pela nVidia (MICHAEL ROMERO, 2011).

O mais alto nıvel deste controle e provido pelos seguintes identificadores: __global__,

__host__ e __device__. Sendo:

• __host__: denota uma funcao que executa na CPU. E o tipo de identificador padrao

e pode fazer tudo que qualquer outra funcao escrita em linguagem de programacao

C++ possa fazer;

• __device__: denota uma funcao que executa na GPU e que nao pode ser invocada

pelo host ;

• __global__: identificador utilizado pelas funcoes do kernel. Estas funcoes podem

ser chamadas do host dentro da device.

Tao importante quanto os identificadores de funcao, sao as variaveis que armazenam

os indices de execucao das threads. Sao elas:

18

• threadIdx: variavel do tipo dim3 7, e utilizada para armazenar as threads dentro

de um thread block ;

• blockIdx: variavel do tipo dim3 que armazena os ındices dos thread blocks dentro

da grid.

Juntamente com a primitiva __syncthreads() estas variaveis sao responsaveis por

garantir a sincronia das threads (MICHAEL ROMERO, 2011).

A figura 2.9 representa a sequencia de execucao de um aplicativo escrito para a pla-

taforma CUDA. Nela tambem podemos notar a divisao da execucao, onde o codigo mas-

sivamente paralelo executa atraves da device e o host se encarrega da execucao do codigo

sequencial/serial.

7Um vetor do tipo uint3, e utilizado para especificar dimensoes na arquitetura CUDA (NVIDIA, 2010b).

19

Figura 2.9: Sequencia de execucao do aplicativo entre a devide e o host


20

2.7.3.2 Kernels e hierarquia de Threads

CUDA C estende a possibilidade de criacao de funcoes existentes na linguagem C.

Estas funcoes sao chamadas de kernels e quando invocadas, podem ser executadas para-

lelamente N vezes por N threads CUDA diferentes (NVIDIA, 2010b).

Um kernel e definido pela utilizacao do identificador __global__ e durante a sua

chamada deve-se especificar o numero de threads CUDA que irao executar este kernel

especificando o configurador de execucao <<<gridSize, blockSize>>>. No quadro 2.1

pode-se ver como e feita a chamada de um kernel.

Quadro 2.1: Exemplo da chamada de um kernel

1 nomeDoKernel <<<gridDim , blockDim>>>(@parametros ) ;

Para cada thread que executa um kernel e dado um identificador unico de execucao,

este entao fica armazenado em um variavel built-in8 chamada threadIdx.

Ikeda (2011) afirma que este modelo de divisao estimula a visao do problema em dois

passos:

• Separar os problemas independentes em formato de grids

• Dividir o grid em blocos independentes de mesmo tamanho (chamados de thread

block) onde estas threads trabalham de forma cooperativa e compartilham uma

memoria visıvel somente entre elas.

A figura 2.10, apresenta esta a forma de divisao dos grids em blocks.

8Variaveis Built-in sao as que especificam o tamanho do bloco e do grid de threads. Sa utilizadasdentro da device (NVIDIA, 2010b)

21

Figura 2.10: Grids de thread blocks do modelo de programacao CUDA


2.7.3.3 Hierarquia de Memoria

Existem diferentes nıveis de memoria dentro de uma GPU, cada uma com uma forma

distinta de leitura e escrita (MICHAEL ROMERO, 2011). A tabela 2.2 apresenta os diferentes

tipos de memoria presentes em uma device CUDA e seus respectivos metodos de acesso e

localizacao.

22

Memoria on/off chip Acesso Escopo Tempo de Vida

Registrador on Leitura/Escrita 1 thread thread

Local off Leitura/Escrita 1 thread thread

Compartilhada on Leitura/Escrita Todas as threads do bloco bloco

Global off Leitura/Escrita Todas as threads + host Alocacao do host

Constante off Leitura Todas as threads + host Alocacao do host

Textura off Leitura Todas as threads + host Alocacao do host

Tabela 2.2: Tipos de memoria presentes no device

(IKEDA, 2011)

2.7.3.4 CUDA Compute Capability

Os principais recursos disponıveis em cada GPU sao expostos atraves de um sistema

padrao chamado compute capability, formado por um major number e um minor number

(IKEDA, 2011).

Atraves desta numeracao pode-se determinar a revisao do conjunto de instrucoes su-

portadas pelo dispositivo (NVIDIA, 2010b). A tabela 2.3 apresenta as principais mudancas

em relacao a cada nova versao da compute capability.

23

Tabela 2.3: Principais funcoes suportadas para cada versao de compute capability

Fonte: adaptado de Ikeda (2011)

2.7.3.5 Codigo de Exemplo

O quadro 2.2 apresenta um aplicativo de exemplo que faz a escrita da frase hello world

na saida do console utilizando a API CUDA sobre a liguagem de programacao C++.

24

Quadro 2.2: Codigo de um aplicativo do tipo hello world escrito em CUDA. Fonte: Galbraith

(2009)

1 #inc lude <cuda . h>

2 #inc lude <s t d i o . h>

3

4 // De f i n i cao da funcao

5 g l o b a l void hel loWorld ( char ∗) ;

6

7 // funcao que e x e cu t a ra no ho s t

8 in t main ( i n t argc , char∗∗ argv )

9 {10 in t i ;

11

12 // Saida de s e j ada

13 char s t r [ ] = " H e l l o W o r l d ! " ;

14

15 // Seprar o conteudo da sa i da

16 f o r ( i = 0 ; i < 12 ; i++)

17 s t r [ i ] −= i ;

18

19 // Aloca a memoria no d e v i c e

20 char ∗ d s t r ;

21 s i z e t s i z e = s i z e o f ( s t r ) ;

22 cudaMalloc ( ( void ∗∗)&d s t r , s i z e ) ;

23

24 // Copia a s t r i n g para a d e v i c e

25 cudaMemcpy( d s t r , s t r , s i z e , cudaMemcpyHostToDevice ) ;

26

27 // d e f i n e o tamanho da g r i d e do b l o c k de t h r e ad s

28 dim3 dimGrid (2) ; // um b l o c o por pa l a v r a

29 dim3 dimBlock (6) ; // uma th r ead por c a r a c t e r

30

31 // invoca o k e r n e l

32 helloWorld<<< dimGrid , dimBlock >>>(d s t r ) ;

33

34 // recupera os r e s u l t a d o s da d e v i c e

35 cudaMemcpy( st r , d s t r , s i z e , cudaMemcpyDeviceToHost ) ;

36

37 // l i b e r a a memoria a l ocada na d e v i c e

38 cudaFree ( d s t r ) ;

39

40 // e x i b e na t e l a

41 p r i n t f ( " % s \ n " , s t r ) ;

42

43 return 0 ;

44 }45

46 // k e r n e l u t i l i z a d o na d e v i c e

47 g l o b a l void hel loWorld ( char∗ s t r ) {48

49 // determina em qua l t h r ead estamos

50 in t idx = blockIdx . x ∗ blockDim . x + threadIdx . x ;

51

52 // remonta o r e s u l t a d o

53 s t r [ idx ] += idx ;

54 }

25

3 Ambiente Experimental

O ambiente experimental utilizado baseia-se em ferramentas proprietarias e open-

source, uma placa de vıdeo capaz de executar codigos da plataforma CUDA e um proces-

sador com multiplos nucleos.

3.1 Estrutura Fısica

Nesta secao sera apresentada a estrutura fısica utilizada, o hardware empregado no

ambiente de testes bem como suas respectivas especificacoes.

3.1.1 Maquina Utilizada

O ambiente de testes e constituıdo por uma unica maquina equipada com as seguintes

configuracoes:

• Placa Mae AsRock P67 Extreme6;

• Placa de vıdeo Zotac GeForce GTX580 com clock 1 de 772Mega-hertz (Mhz) e 512

cuda cores de processamento, 1536 Megabytes (MB) de memoria Graphics Double

Data Rate (GDDR)5 e capacidade nominal de processamento para operacoes de

ponto flutuante de 1581.1 Floating-point Operations per Second (GFLOP);

• 2 Pentes de memoria Corsair Vengeance com capacidade de 4 Gigabytes (GB) tota-

lizando 8 GB;

• Processador Intel Core i5 2500K, com 4 nucleos de processamento e clock de 3.3Giga-

hertz (Ghz) e 6MB de cache nıvel 3;

1Denominacao dada ao sinal oscilante utilizado por dispositivos eletronicos como metodo de sincroniapara transmissao de dados. (WIKIPEDIA, 2011a)

26

O uso de uma maquina diferente pode acarretar em diferencas significativas nos re-

sultados.

3.2 Estrutura Logica

Nesta secao serao apresentados os principais softwares e bibliotecas utilizadas na cons-

trucao dos cenarios para os testes.

3.2.1 Sistema Operacional

A linguagem de programacao C++ e a API CUDA sao compatıveis com a maioria

dos sistemas operacionais modernos. Neste trabalho sera utilizado um sistema baseado

em Unix, Ubuntu Linux versao 10.04.3 LTS de 64 bits, para a realizacao dos testes.

3.2.2 Plataforma CUDA

CUDA e uma arquitetura de computacao paralela onde atraves do aproveitamento do

poder das GPU’s tem-se um aumento expressivo no desempenho computacional. (NVIDIA,

2011c)

Neste trabalho sera utilizada a arquitetura CUDA da versao 2.0 por ser a mais nova

suportada pelo hardware disponıvel.

3.2.2.1 CUDA Toolkit

Kit de ferramentas da plataforma CUDA, nele podem ser encontradas as bibliotecas,

compiladores e documentacoes necessarias para a criacao de aplicacoes que utilizem a

plataforma CUDA. (NVIDIA, 2011b)

Neste trabalho sera utilizada a versao 3.2 do CUDA Toolkit para sistemas de 64 bits.

3.2.3 3Ds MAX 2012

O 3Ds Max e um ambiente integrado de desenvolvimento para modelagem, animacao

e renderizacao tridimensional capaz de proporcionar aos seus usuarios um experiencia de

criacao muito agil e rapida. (AUTODESK, 2011)

27

Para a execucao dos testes de desempenho entre as plataformas, este trabalho ira

utilizar um objeto tridimensional criado no 3Ds Max 2012 e exportado para o padrao

de arquivo .3ds que e reconhecido pela renderer customizada utilizada para comparar as

plataformas GPGPU e em CPU.

3.2.4 Renderer C++

Renderer que utiliza uma implementacao de algoritmos baseado em polıgonos 3D.

Suas principais caracterısticas sao a presenca de um Raytracer e processamento paralelo

do calculo de iluminacao dos pixels atraves da API Open Multi-Processing (OpenMP)2

ou Intel Threading Building Blocks (Intel TBB). (TSIODRAS, 2010)

Este trabalho utilizara a versao 2.2s da renderer, sendo esta a ultima versao estavel do

projeto. Nela serao executadas as medicoes de desempenho referente a CPU multinuclear.

3.2.4.1 Simple DirectMedia Layer

O Simple DirectMedia Layer (SDL) e uma biblioteca multimıdia escrita em lingua-

gem C desenhada para oferecer acesso em baixo nıvel a interacao com teclado, audio,

framebuffer de vıdeo 2D, 3D via OpenGL, joysticks e mouse. (LANTINGA, 2011)

Este trabalho utilizara como base para geracao dos graficos processados na CPU a

biblioteca SDL versao 1.2.

3.2.5 Renderer CUDA

Uma modificacao da renderer escrita na linguagem C++ para a linguagem CUDA

possibilitando entao o uso de uma GPGPU como processador paralelo para os calculos

referentes a iluminacao. (TSIODRAS, 2011)

Este trabalho ira utilizar a versao 2.2h desta, que tambem sera utilizada para as

medicoes de desempenho da plataforma GPGPU.

3.2.5.1 OpenGL

O OpenGL e a biblioteca especificada como padrao para escrita de aplicativos que

utilizem graficos em tres ou duas dimensoes, por ser multiplataforma pode ser utilizado

2Uma API para oferecer suporte a programacao paralela com memoria compartilhada em multiplasplataformas atraves das linguagens de programacao C/C++ ou Fortran

28

em qualquer tido de sistema. E mantido pelo consorcio Khronos Group como uma API

Open-Source. (KHRONOS, 2011b)

Para a exibicao das imagens renderizadas pela renderer CUDA sera utilizada a bibli-

oteca de graficos tridimensionais OpenGL 1.4.

3.2.6 Plataforma C++

Nesta secao serao descritas as bibliotecas e compiladores utilizados para a compilacao

e execucao da renderer escrita sobre a linguagem C++.

3.2.6.1 Compilador

O GNU Compiler Collection (GCC), compilador de varias linguagens produzido pelo

projeto GNU (refere-se ao animal Gnu que tambem e o mascote da organizacao) e distri-

buıdo pela Free Software Foundation. E um componente essencial para indispensavel para

sistemas baseados em Unix (como Linux) sendo em algumas distribuicoes o compilador

padrao do sistema. (WIKIPEDIA, 2011b)

Possui amplo suporte a linguagem de programacao C e atraves de extensoes pode-se

adicionar suporte a outras linguagens como C++, FORTRAN e JAVA.

Sera utilizado para a compilacao da renderer baseada na linguagem C++ o compilador

GCC de 64 bits versao 4.4.3 com extensoes para linguagem C++.

3.2.6.2 Intel R© Threading Building Blocks

Uma rica API criada pela Intel Corporation para expressar paralelismo em aplicativos

programados utilizando a linguagem C++ sem que seja necessario ser um especialista em

threads. (INTEL, 2011)

Neste trabalho sera utilizada a versao 4.0 da biblioteca Intel TBB para sistemas ba-

seados em CPU’s multinucleares.

29

4 Implementacao

Esta secao visa explorar os detalhes da implementacao e realizacao dos testes sobre

a plataforma proposta com o objetivo de avaliar o tempo gasto para a renderizacao dos

modelos tridimensionais e a taxa de exibicao das imagens geradas na tela.

4.1 Descricao do Estudo de Caso

O estudo de caso utilizado por este trabalho contempla uma avaliacao do tempo, em

segundos, necessario para o calculo das interseccoes de raios do ray tracer no modelo

tridimensional e a velocidade de apresentacao da imagem na tela em FPS.

Para tais demonstracoes foram utilizados tres modelos tridimensionais diferentes.

Cada modelo apresenta uma complexidade diferente em relacao a quantidade de trian-

gulos e vertices, para este trabalho foram utilizados tres nıveis de complexidade: baixa,

media e alta.

Descricao dos modelos:

Chessboard : um tabuleiro de xadrez. E o modelo mais simples dos tres utilizados

possui 46658 triangulos 32488 vertices, pode ser visto na figura 4.11;

Dodge Challenger 2007 : um carro esportivo americano, colorido com preto fosco e

vidros verdes. E modelo de complexidade mediana, possui 1067082 triangulos e

355694 vertices, pode ser visto na figura 4.12;

Happy : um buda contente. E o modelo mais complexo utilizado nos testes, pois possui

543652 triangulos 1087716 vertices, pode ser visto na imagem 4.13

30

Figura 4.11: Modelo de complexidade simples, Chessboard

Fonte: Tsiodras (2010)

Figura 4.12: Modelo de complexidade media, Dodge Challenger 2007

Fonte: Gregorio (2011)

31

Figura 4.13: Modelo de complexidade alta, Happy

Fonte: Diaz (2011)

4.1.1 Descricao dos Cenarios de Testes

Para os testes de desempenho foram utilizados dois cenarios diferentes para cada

modelo em cada uma das plataformas. Em cada cenario foi medida a velocidade do

calculo de renderizacao em segundos e a velocidade de exibicao da cena quadro-a-quadro

pela quantidade de FPS.

4.1.1.1 Cenarios

Em ambas as plataformas foram utilizadas os seguintes modos de renderizacao para

os modelos tridimensionais e obtencao dos dados:

Modo 9 : renderizacao do modelo tridimensional e exibicao de 60 quadros utilizando

efeito de iluminacao por raytracing sem suavizacao das bordas com o uso de anti-

aliasing1.

Modo 0 : renderizacao do modelo tridimensional e exibicao de 60 quadros utilizando

1Tecnica que permite suavizar os contornos de imagens adicionando pontos de cores intermediarios assuas bordas (MORIMOTO, 2006).

32

efeitos de iluminacao por ray tracing e suavizacao de bordas das imagens utilizando

anti-aliasing Morphological Anti Aliasing (MLAA);

4.2 Resultados

Nesta secao serao apresentados os resultados obtidos em cada teste para as plata-

formas GPGPU utilizando CUDA e para a plataforma de CPU multinuclear utilizando

a linguagem C++ com paralelismo implementado via TBB em cada um dos modos de

renderizacao apresentados.

4.2.1 Modelo de baixa complexidade, Chessboard

Como pode ser visto na tabela 4.4, a plataforma CUDA foi mais rapida em relacao a

CPU em ambos os modos de renderizacao.

Tabela 4.4: Resultados obtidos, modelo de baixa complexidade

Como pode ser visto no grafico 4.14, quando nao utilizado anti-aliasing a diferenca

de desempenho no tempo de renderizacao foi de 24.515 segundos, ou seja, a GPGPU

renderizou o modelo tridimensional em 2,38% do tempo levado pela CPU.

33

Figura 4.14: Tempo de renderizacao em cada cenario para o modelo de baixa complexidade

O grafico 4.15 apresenta a diferenca na taxa de exibicao de quadros na tela, tambem

mais alta quando utilizado o sistema GPGPU.

Figura 4.15: Taxa de FPS para o modelo de baixa complexidade

4.2.2 Modelo de media complexidade, Dodge Challenger 2007

Com o aumento na complexidade do modelo, houve uma perda consideravel de desem-

penho em ambas a plataformas quando utilizado anti-aliasing, porem o CUDA continua

sendo mais rapido que a CPU multinuclear para renderizacao e exibicao das cenas como

34

mostrado na tabela 4.5.

Tabela 4.5: Resultados obtidos, modelo de media complexidade

O sistema GPGPU realizou o processamento do modelo em modo 0 com apenas 12,24%

do tempo gasto pela CPU, como pode ser visto no grafico 4.16.

Figura 4.16: Tempo de renderizacao em cada cenario para o modelo de media complexidade

A taxa de exibicao na tela tambem se manteve mais alta para o sistema GPGPU em

relacao a CPU, como pode ser visto no grafico 4.15.

35

Figura 4.17: Taxa de FPS para o modelo de media complexidade

4.2.3 Modelo de alta complexidade, Happy

O modelo de maior complexidade utilizado neste trabalho, com mais de 1 milhao

de vertices. Este modelo foi renderizado com sucesso na CPU multinuclear, entretanto

o sistema GPGPU devido a um limitacao no tempo limite de execucao de processos da

plataforma CUDA nao conseguiu completar o processamento sem que a maquina utilizada

congelasse.

A tabela 4.6 apresenta os resultados obtidos com a CPU, e como pode ser visto,

devido a complexidade do modelo foram resultados com valores muito baixos em relacao

aos outros modelos previamente utilizados.

Tabela 4.6: Resultados obtidos, modelo de alta complexidade

36

O grafico 4.18 apresenta o tempo gasto para renderizacao do modelo. Se comparado

ao tempo gasto pela CPU durante a renderizacao do modelo Dodge Challenger, houve

um aumento de mais de 30%.

Figura 4.18: Tempo de renderizacao em cada cenario para o modelo de alta complexidade

Com taxa de FPS abaixo de 1 no modo de renderizacao 0, ao alterar a renderizacao

para o modo 9 houve um ganho de 0.85 FPS durante a exibicao da imagem, como pode

ser visto no grafico 4.19.

Figura 4.19: Taxa de FPS para o modelo de alta complexidade

37

5 Consideracoes Finais

Neste capıtulo estao descritos a conclusao da pesquisa e trabalhos futuros que se

possam realizar.

5.1 Problemas Durante os Testes

Durante a realizacao do ultimo teste de desempenho, utilizando o modelo Happy,

o aplicativo de renderizacao que utiliza a API CUDA para paralelismo, nao conseguiu

terminar o processamento das insercoes dos raios de iluminacao por ray tracing no modelo

devido ao processamento ter excedido o tempo maximo de execucao para um placa de

vıdeo em modo compartilhado (modo em que a device alem de realizar o processamento

do aplicativo CUDA, realiza o processamento do ambiente grafico do sistema operacional)

no Linux.

De acordo nVidia (2010a) o CUDA Toolkit 3.2 que foi utilizado neste trabalho, possui

uma limitacao para GPU’s em modo compartilhado, o tempo de execucao maximo para

aplicativos que utilizam a API CUDA e limitado a 5 segundos. Este tempo ainda pode

ser acrescido de 1 ou 2 segundos dependendo da capacidade da placa.

Levando em conta a utilizacao de uma GTX580, placa mais potente da serie GeForce

disponıvel no mercado com apenas uma GPU, os tempos indicados sao coerentes, visto

que o modelo Dodge Challenger foi processado com sucesso em 6 segundos.

Finalizando o ambiente grafico atraves do comando /etc/init.d/gdm stop e exe-

cutando novamente o aplicativo de renderizacao, foi possıvel notar que o processo de

interseccao dos raios foi realizado porem sem o ambiente de exibicao em janelas nao foi

possıvel exibir a imagem processada nem mensurar os dados de tempo de renderizacao e

FPS.

Tendo em vista o resultado obtido finalizando o ambiente grafico, fica claro que a

utilizacao de uma unica GPU para processamento pesado nao e possıvel em ambiente de

38

producao, sendo necessaria a utilizacao de duas GPU’s, uma para o processamento do

ambiente grafico e outra para o processamento dos aplicativos CUDA.

5.2 Conclusao

Este trabalho de conclusao de curso apresentou uma comparacao de desempenho

entre as plataformas GPGPU e CPU multinuclear para processamento paralelo utilizando

modelos tridimensionais iluminados por ray tracing finalizando com a realizacao de testes

e apresentacao dos resultados obtidos com base em cada cenario apresentado.

Com os testes aqui apresentados foi possıvel comparar ambas as plataformas e concluir

que, a plataforma GPGPU se mostrou em media 1800% mais rapida que a CPU multi-

nuclear para processamentos que envolvem alto paralelismo de dados, entretanto deve-se

observar que para um cenario de producao onde nao devera haver paradas por exceder o

tempo de processamento e necessaria a utilizacao de uma placa dedicada para execucao

dos aplicativos que utilizem API CUDA para paralelismo.

5.3 Trabalhos Futuros

Com base nos resultados obtidos utilizando-se do processamento de uma placa da serie

GeForce, criar um novo estudo de caso utilizando uma placa especıfica para a execucao de

aplicativos CUDA, como por exemplo uma placa da serie Tesla1, possivelmente obtendo

melhores resultados em relacao aos apresentados.

1GPU’s desenvolvidas para computacao de alto desempenho capazes de oferecer ate 10x mais veloci-dade para os aplicativos CUDA (NVIDIA, 2011a).

39

Referencias

ABICHAHLA, F. The CUDA APIs. 2008. Disponıvel em:<http://www.tomshardware.com/reviews/nvidiacudagpu,19546.html>. Acessoem: 21/10/2011.

ALVARO CUNO, J. R. M. V. Conceitos de programacao em GPU. 2011. Disponıvel em:<http://www.visual.pro.br/dl9>. Acesso em: 17/10/2011.

AUTODESK, I. Produtos Autodesk 3ds Max. 2011. Disponıvel em:<http://www.autodesk.com.br/adsk/servlet/pc/index?id=14644213&siteID=1003425&mktvar004=361323>.Acesso em: 07/10/2011.

BARNEY, B. Introduction to Parallel Computing. 2011. Disponıvel em:<https://computing.llnl.gov/tutorials/parallel comp/>. Acesso em: 20/10/2011.

BARTLETT, J. Programming from the Ground Up. [S.l.]: Bartlett Publishing, 2003.

BIRN, J. 3D Rendering. 2002. Disponıvel em:<http://www.3drender.com/glossary/3drendering.htm)>. Acesso em: 19/10/2011.

BRAIN, M. How Microprocessors Work. 2000. Disponıvel em:<http://computer.howstuffworks.com/microprocessor.htm>. Acesso em: 16/10/2011.

DAVID B. KIRK, W.-M. W. H. Programando para Processadores Paralelos, UmaAbordagem Pratica a Programacao de GPU. Rio de Janeiro: Campus / Elsevier, 2010.

DIAZ, P. starting point. 2011. Disponıvel em:<http://www.ics.uci.edu/ pablo/files/data/genusnon0/happy.ply>. Acesso em:28/11/2011.

FARIA, A. GPU e CPU o futuro com computacao acelerada. 2009. Disponıvel em:<http://www.gentefina.com.br/materia/Alessandro%20Faria/65/72/>. Acesso em:16/10/2011.

GALBRAITH, B. Hello World using CUDA. 2009. Disponıvel em:<http://forums.nvidia.com/index.php?showtopic=90044>. Acesso em: 21/10/2011.

GREGoRIO, R. Dodge Challenger 2007. 2011.

GUSMaO, F. Introducao a programacao paralela em sistemas heterogeneos com OpenCL.2010. Disponıvel em: <http://blogdogusmao.wordpress.com/2010/01/21/introducao-opencl/>. Acesso em: 16/10/2011.

HARDWARE CLUBE do. Versoes do DirectX. 2008. Disponıvel em:<http://www.clubedohardware.com.br/artigos/964>. Acesso em: 17/10/2011.

40

HORSTMANN, C. Padroes e Projeto Orientados a Objetos. [S.l.]: Bookmark, 2006.

IKEDA, P. A. Um Estudo do uso Eficiente de Programas em Placas Graficas. 2011.Disponıvel em: <http://gpubrasil.files.wordpress.com/2011/09/dissertacao-versao-entregue.pdf>. Acesso em: 18/10/2011.

INTEL, C. Intel Threading Building Blocks for OpenSource. 2011. Disponıvel em:<http://threadingbuildingblocks.org/>. Acesso em: 13/10/2011.

JOHN D. OWENS MIKE HOUSTON, D. L. S. G. J. E. S. J. C. P. GPU Computing. 2008.Disponıvel em: <http://cs.utsa.edu/qitian/seminar/Spring11/03 04 11/GPU.pdf>.Acesso em: 19/10/2011.

KHRONOS, G. OpenCL. 2011. Disponıvel em: <http://www.khronos.org/opencl/>.Acesso em: 13/10/2011.

KHRONOS, G. OpenGL. 2011. Disponıvel em: <http://www.khronos.org/opengl>.Acesso em: 13/10/2011.

LANTINGA, S. SDL: Simple Directmedia Layer. 2011. Disponıvel em:<http://www.libsdl.org>. Acesso em: 13/10/2011.

LIN, C. Fetch, Execute, Decode. 2003. Disponıvel em:<http://www.cs.umd.edu/class/sum2003/cmsc311/Notes/CompOrg/fetchDecode.html>.Acesso em: 09/11/2011.

MICHAEL ROMERO, R. U. CUDA Programming. 2011. Disponıvel em:<http://cuda.ce.rit.edu/cuda overview/cuda overview.htm>. Acesso em: 21/10/2011.

MORIMOTO, C. E. AntiAliasing. 2006. Disponıvel em:<http://www.hardware.com.br/termos/antialiasing>. Acesso em: 12/11/2011.

NVIDIA, C. CUDA Toolkit 3.2 Release Notes for Linux. 2010. Disponıvel em:<http://developer.download.nvidia.com/compute/cuda/3 2/toolkit/docs/CUDA Toolkit Release Notes Linux.txt>.Acesso em: 28/11/2011.

NVIDIA, C. NVIDIA CUDA C Programming Guidev3.2. nVidia Corporation, 2010. Disponıvel em:<http://developer.download.nvidia.com/compute/cuda/3 2/toolkit/docs/CUDA C Programming Guide.pdf>.Acesso em: 20/10/2011.

NVIDIA, C. Computacao de Alta Performance. 2011. Disponıvel em:<http://www.nvidia.com.br/object/tesla computing solutions br.html>. Acessoem: 28/11/2011.

NVIDIA, C. CUDA Downloads. 2011. Disponıvel em:<http://developer.nvidia.com/cudadownloads>. Acesso em: 07/10/2011.

NVIDIA, C. What Is CUDA? 2011. Disponıvel em: <http://developer.nvidia.com/what-cuda>. Acesso em: 07/10/2011.

OWEN, G. S. Ray Tracing. 1999. Disponıvel em:<http://www.siggraph.org/education/materials/HyperGraph/raytrace/rtrace0.htm>.Acesso em: 19/10/2011.

41

RITA BORGO, K. B. State of the Art Report on GPU Visualization. 2009. Disponıvelem: <http://www.viznet.ac.uk/files/VIZNETWP424LEEDSGPUweb.pdf>. Acesso em:20/10/2011.

SANTEE, A. Programacao de Jogos com C++ e DirectX. [S.l.]: Novatec Editora, 2005.

SCHETINGER, V. CUDA Compute Unified Device Architecture. 2009. Disponıvel em:<http://wwwpet.inf.ufsm.br/periodico/index.php/tutoriais/6mensais/51cudacompute-unifieddevicearchitecture>. Acesso em: 20/10/2011.

SILVA, F. W. S. V. da. Introducao ao Ray Tracing. 2011. Disponıvel em:<http://www.visgraf.impa.br/Data/RefBib/PS PDF/silva94/rt.pdf>. Acesso em:19/10/2011.

SILVA BASTOS SALES, P. da. Avaliacao de Desempenho de Ferramentas deRenderizacao de Imagens em Clusters openMosix e Arquiteturas Multicore. 2008.Disponıvel em: <http://dsc.upe.br/tcc/20081/monografiaPericlesvf.pdf>. Acesso em:09/11/2011.

SIPAHI, P. D. G. Coloquio Computacao Heterogenea: dos Videogames para aFısica Computacional. 2009. Disponıvel em: <http://boletim.ifsc.usp.br/Todas-Noticias.php?rowid home=1604&rowid vol=80>. Acesso em: 17/10/2011.

SOUZA, E. T. T. P. Introducao a Computacao de Alto Desempenho Utilizando GPU.2011. Disponıvel em: <http://gpubrasil.files.wordpress.com/2011/04/introducaoa-computacaodealtodesempenhoutilizandogpu1.pdf>. Acesso em: 18/10/2011.

TANENBAUM, A. S. Organizacao Estruturada de Computadores. [S.l.]: Prentice-Hall,2006.

TOP500. Segments share for 06/2011. 2011. Disponıvel em:<http://top500.org/charts/list/37/segments>. Acesso em: 20/10/2011.

TSIODRAS, D.-I. T. A realtime 3D renderer in OpenMP/TBB. 2010. Disponıvel em:<http://users.softlab.ece.ntua.gr/ ttsiod/renderer.html>. Acesso em: 07/10/2011.

TSIODRAS, D.-I. T. A realtime raytracer of triangle meshes in CUDA. 2011. Disponıvelem: <http://users.softlab.ece.ntua.gr/ ttsiod/cudarendererBVH.html>. Acesso em:07/10/2011.

VIANA, J. R. M. Programacao em GPU: Passado, Presente e Futuro. 2009. Disponıvelem: <http://www.ufpi.br/subsiteFiles/ercemapi/arquivos/files/minicurso/smc1.pdf>.Acesso em: 16/10/2011.

WHITTED, T. An Improved Illumination Model for Shaded Display. 1980. Disponıvel em:<http://citeseerx.ist.psu.edu/viewdoc/download?doi=10.1.1.114.7629&rep=rep1&type=pdf>.Acesso em: 19/10/2011.

WIKIPEDIA. Clock Signal. 2011. Disponıvel em:<http://en.wikipedia.org/wiki/Clock signal>. Acesso em: 13/10/2011.

WIKIPEDIA. GNU Compiler Collection. 2011. Disponıvel em:<http://pt.wikipedia.org/wiki/GNU Compiler Collection>. Acesso em: 07/10/2011.

42

WIKIPEDIA. Microsoft Direct3D. 2011. Disponıvel em:<http://en.wikipedia.org/wiki/Microsoft Direct3D>. Acesso em: 17/10/2011.

WIKIPEDIA. Raster graphics. 2011. Disponıvel em:<http://en.wikipedia.org/wiki/Raster graphics)>. Acesso em: 19/10/2011.

WIKIPEDIA. Ray Tracing (graphics). 2011. Disponıvel em:<http://en.wikipedia.org/wiki/Ray tracing (graphics)>. Acesso em: 19/10/2011.

ZIBULA, A. General Purpose Computation on Graphics ProcessingUnits (GPGPU) using CUDA. 2009. Disponıvel em: <http://www.wi.uni-muenster.de/pi/lehre/ws0910/pppa/papers/gpgpu.pdf>. Acesso em: 19/10/2011.

Documents

GPU Computing vs. CPU Multinuclear: Análise de desempenho para processamento de iluminação por Ray Tracing em modelos tridimensionais