Lista de questões para a prova de RV (responda todas as questões baseando-se no programming guide disponibilizado com os exercícios práticos) 1) Fale sobre como são organizadas as threads, blocos e grids em CUDA. 2) Compare as arquiteturas SIMT (CUDA) e SIMD (Intel SSE). 3) Como funciona e para que serve a função __synchthreads()? A função __syncthreads() serve para, dentro de um kernel, estabelecer um ponto de sincronização. Assim as threads daquele bloco só podem passar o ponto de sincronização se todas as outras estiverem prontas. Ela é útil para se trabalhar com memória compartilhada, em que as threads esperam as outras terminarem suas operações na 'shared memory' para então prosseguir com a execução do kernel. 4) Para que serve a função cudaThreadSynchronize()? Como a chamada de um kernel não bloqueia a CPU, esta função se faz necessária caso as próximas instruções da CPU dependam do resultado das operações feitas em GPU (o que é verdade na grande maioria dos casos). Portanto a função cudaThreadSynchronize() bloqueia o processo em CPU enquanto o kernel não termina de ser executado. 5) Qual o número máximo de threads que se pode ter em uma configuração de um bloco (para device capability 2.x)? Em GPUs atuais, um bloco pode conter até 1024 threads[1] [1] CUDA C Programming Guide Version 3.2, p8 -6) Compare as seguintes memórias (tempo de acesso e espaço): constante, textura, global, local e compartilhada. A memória local é privada para cada thread e, dentre os tipos de memória citados, é o que tem o tempo de acesso mais rápido. Cada bloco de threads tem também uma memória compartilhada que pode ser acessada por todas as threads do bloco. A memória compartilhada é mais rápida que a global. As memórias global, constante e de textura são visíveis por todas as threads, independente do bloco. 7) Qual a quantidade máxima em KBytes de memória compartilhada que pode ser utilizada em um bloco? Nos devices 1.x, cada bloco tem somente 16KB de espaço de memória compartilhada. 8) Liste 4 modelos diferentes de placas de vídeo com suporte à CUDA em ordem de desempenho. GeForce GTX 460 GeForce GTX 295 GeForce GT 330M GeForce 8600 GT 9) Qual o número de threads em um warp? Nos devices 1.x um warp contem 32 threads 10) Cite 3 diferentes “built-in vector types” de CUDA. char2, float2 e int2 11) Em um filtro de média em CUDA, qual a vantagem de se utilizar textura 2D? ???O acesso a uma estrutura 2D fica mais fácil??? Texturas são otimizadas para trabalhar em um espaço 2D. (p93) 12) Como se acessa o terceiro elemento de um tipo vetorial de CUDA com 4 elementos? Através do campo .z Exemplo: Na definição float4 vetor = make_float4(1.0f,2.0f,3.0f,4.0f) o terceiro elemento pode ser acessado da forma vetor.z 13) Qual o retorno da função make_float4(1.0f,2.0f,3.0f,4.0f)? Um vector do tipo float4 com o valor (1.0f,2.0f,3.0f,4.0f) 14) Para que serve o tipo dim3? O dim3 é um vector do tipo inteiro baseado no uint3. Ele é usado para especificar dimensões como por exemplo as dimensões do bloco na chamada de um kernel. ?15) Para que servem as operações de “bind” e “unbind” em texturas? O bind serve para dar acesso ao kernel ao espaço de memória usado pela textura. 16) Qual a função que se utiliza para alocar um espaço de memória na GPU? Qual a função utilizada para desalocar este mesmo espaço de memória? Para alocar utilizamos a função cudaMalloc(**devPtr, size), que recebe a quantidade de bytes que serão alocados e uma variável que reberá o endereço do espaço de memória alocado. Para liberar esse mesmo espaço utiliza-se a função cudaFree() 17) Qual a função utilizada para acessar um elemento de uma textura unidimensional? tex1Dfetch(texture, index) 18) Explique as diferenças entre os qualificadores: __host__, __device__ e __global__. Tomando como exemplo a declaração de funcções: __host__ qualifica uma função executada no host que só pode ser chamada por outras funções do host __global__ qualifica uma função que é executada no device, mas pode ser chamada tanto pelo host quanto pelo device __device__ indica que a função só poderá ser chamada por outtra função do device 19) Qual qualificador de CUDA vem sempre no início da declaração do código de um kernel? O que ele indica? O qualificador __global__ é usado e ele indica que apesar da função ser executada em GPU, ela também pode ser chamada a partir do host. Isso permite que a CPU possa chamar o kernel. 20) O que significa cada um dos parâmetros que são passados dentro de uma chamada de kernel, entre “<<<” e “>>>”? Na maioria dos casos são passados apenas 2 parâmetros: o primeiro indica as dimensões do grid (ou quantos blocos devem existir no grid) e o segundo indica as dimensões de cada bloco do grid (ou quantas threads existirão por blocos). Apesar disso existem mais 2 parametros. Um deles é o número de bytes alocados dinamicamente na memoria compartilhada. 21) Quando é a vantagem em usar memória compartilhada? Quando não é vantagem utilizá-la? É aconselhado usar memória compartilhada nos casos em que mais de uma leitura na memória global é necessária. 22) Explique o que são conflitos de banco, por que eles ocorrem e como fazer para evitá-los. 23) Explique o que é um código PTX. 24) Quando acontece o caso de broadcast na memória compartilhada? 25) O que faz o seguinte trecho de código em PTX? ld.global.v4.f32 {$f1,$f3,$f5,$f7}, [$r9+0]; mad.f32 $f1, $f3, $f5, $f1; 26) Para que serve o CUDA Profiler? Quais (cite algumas) informações sobre a aplicação é possível conseguir com seu uso? 27) Qual técnica de programação pode ser utilizada para diminuir o número de “branches” de uma aplicação? 28) Suponha que você deve implementar uma aplicação de binarização de uma imagem de tamanho 64x64. Você recomendaria a utilização de CUDA para tal tarefa? Justifique. A imagem é muito pequena e só o overhead de transferir os dados do host para o device pode não compensar a execução de uma operação tão simples como a binarização. -29) Cite 2 exemplos de aplicações que podem aproveitar o potencial da arquitetura fornecida por CUDA. Justifique. Jogos. Possui vários objetos em cena que são "animados" em tempo real. Cada operação feita nos objetos significa multiplicação de várias matrizes e fazer tais operações em paralelo ... 30) Suponha que você deve executar três kernels distintos, com os respectivos tempos de execução: 1, 2 e 3 ms. Qual o tempo total de execução (aproximado)? Cerca de 3ms. Como a execução de um kernel pode ser feita em paralelo com os outros, o tempo total se aproxima do tempo de execução do kernel mais demorado. 31) Qual a vantagem do acesso sequencial sobre o acesso aleatório em memória global? 32) Qual a vantagem do acesso sequencial sobre o acesso aleatório em texturas? 33) Suponha que você deve implementar uma aplicação de binarização de uma imagem de tamanho 4096x4096. Você recomendaria a utilização de CUDA para tal tarefa? Justifique. 34) Como fazer para alocar espaço de memória não-paginada no host? Existe alguma limitação para a quantidade de memória alocada? A limitação depende da CPU (não convém usar todo o espaço de memória não paginada e deixar o sistema operacional sem nada). 35) Como funciona a técnica de “loop unrolling”? Consiste em substituir a definição de um loop pelas linhas de código que seriam executadas em cada ciclo do loop. Isso evitaria que a GPU tivesse que testar a cada ciclo se o loop continuaria sendo executado. 36) Existe diferença entre usar “1.0” e “1.0f” dentro de um código de CUDA? Por quê? 37) Qual seria um possível problema gerado pelo uso de um array da estrutura declarada a seguir e armazenada em memória compartilhada? Justifique. struct { char num; short weight; }; 38) Explique o que significa “CUDA occupancy”. 39) O que acontece se o campo “ReadMode” da struct textureReference for setado para “cudaReadModeNormalizedFloat”? 40) O que acontece se o campo “ReadMode” da struct textureReference for setado para “cudaReadModeElementType”? 41) Para que serve o cudaChannelFormatDesc ? Dê um exemplo de utilização. 42) Em GPUs, por que operações com ponto flutuante são geralmente mais rápidas que operações de inteiro? 43) Porque deve-se priorizar a escrita sequencial em memória global? O que muda em relação à escrita não sequencial? Tempo? Largura de banda? Explique o porquê das mudanças. 44) O que são bancos de memória? Quantos bits são endereçados com um banco? Como podem ser detectados conflitos de banco? 45) Assinale a alternativa correta. Conflitos de banco podem ocorrer entre: a. operações de store/load em diferentes trechos de código. b. operações de store/load num mesmo bloco. c. operações de store/load num mesmo half-warp. Explique sua resposta. ?46) Onde são armazenados os parâmetros dos kernels? Na memória global. ?47) Em um trecho de código dentro de um kernel são utilizadas operações de multiplicação de números inteiros e módulo. Observou-se que todas as operações utilizam do lado direito do operador, constantes numéricas que são potência de 2. Que alteração devemos realizar no código para otimizar estas operações? Shift left 48) Porque não se deve utilizar __syncthreads() dentro de desvios condicionais? 49) Cite 2 casos de uma implementação em CPU ter melhor performance que uma em GPU. Casos em que se faz poucas operações matemáticas e é necessário o uso de muitas estruturas de controle condicionais. Quando o paralelismo é pouco. 50) Para que servem as chamadas operações atômicas? Servem nos casos em que é necessário garantir que os valores trabalhados não serão modificados por outras threads durante a execução de certo trecho de código.