Cuidados na medição de tempo

Adaptado da mensagem enviada por Daniel Medina à lista GPUBrasil.

Pessoal, já faz algum tempo que estou para reviver este tópico e sempre esqueço. Quero compartilhar com vocês um detalhe que me consumiu alguns dias quebrando a cabeça em cima disso.

Ao implementar em meu programa o algoritmo sugerido pelo Labaki para realizar a análise do tempo de execução, notei que o desempenho da GPU estava muito ruim comparado ao da CPU (com volume pequeno de dados estava até maior). Após procurar muito a causa, descobri que o vilão da história é a API do CUDA. Esta camada de software não é inicializada durante a inicialização do programa principal. Ela só é instanciada no momento da primeira chamada a alguma função da API (no meu caso ocorria quando eu chamava a função cuda_malloc pela primeira vez).

# include <stdlib.h>

# include <stdio.h>

# include <cutil.h>

int main()

{

float tempo;

unsigned int timer = 0;

CUT_SAFE_CALL( cutCreateTimer( &timer));

CUT_SAFE_CALL( cutStartTimer( timer));

// Primeiro comando para a GPU (cudaMalloc, por exemplo).

// …

// Região de execução: aqui vai o trecho de código cujo

// tempo de execução queremos medir.

tempo = CUT_SAFE_CALL( cutGetTimerValue(timer));

printf(“A execução durou %5.3f milissegundos\n”, tempo);

CUT_SAFE_CALL(cutDeleteTimer(timer));

}

Ao receber a primeira instrução, a API se inicializa e isto leva um tempo considerável. Empiricamente levantei que esse tempo é de aproximadamente 40ms em uma GTX  280 e 70ms em uma GT 8800.

Para aplicações que duram dias ou semanas, este tempo não chega a ser um problema, mas para aplicações como a minha, cujo tempo de execução é da ordem de milissegundos, este delay é algo altamente indesejável.

Como solução, eu coloquei uma instrução “dummy” (malloc de 0 bytes) no início do algoritmo só para inicializar a placa, e desconsiderei este tempo do cálculo utilizado para comparação entre CPU e GPU. Por ser um custo de tempo intrínseco e que acontece somente uma vez em toda a execução do software, ele pode ser desconsiderado do benchmark. Veja um exemplo.

# include <stdlib.h>

# include <stdio.h>

# include <cutil.h>

int main()

{

float tempo;

unsigned int timer = 0;

CUT_SAFE_CALL( cutCreateTimer( &timer));

// Um comando dummy, sem efeito no programa, só para

// inicializar a placa. Esse passa a ser então o primeiro

// comando de GPU do programa.

float* dummy;

CUDA_SAFE_CALL( cudaMalloc((void**) &dummy, sizeof(float)));

// Só então o cronômetro é ligado:

CUT_SAFE_CALL( cutStartTimer( timer));

// Região de execução: aqui vai o trecho de código cujo

// tempo de execução queremos medir.

tempo = CUT_SAFE_CALL( cutGetTimerValue(timer));

printf(“A execução durou %5.3f milissegundos\n”, tempo);

CUT_SAFE_CALL(cutDeleteTimer(timer));

}

Parte ou todo esse roteiro não funcionou pra você? Deixe comentário!