Тест производительности CUDA

Question

Я пишу простую программу CUDA для теста производительности.
Это не связано с векторным вычислением, а просто для простого (параллельного) преобразования строк.

#include <stdio.h> 
#include <string.h> 
#include <cuda_runtime.h> 


#define UCHAR   unsigned char 
#define UINT32   unsigned long int 

#define CTX_SIZE  sizeof(aes_context) 
#define DOCU_SIZE  4096 
#define TOTAL   100000 
#define BBLOCK_SIZE  500 


UCHAR   pH_TXT[DOCU_SIZE * TOTAL]; 
UCHAR   pH_ENC[DOCU_SIZE * TOTAL]; 
UCHAR*   pD_TXT; 
UCHAR*   pD_ENC; 


__global__ 
void TEST_Encode(UCHAR *a_input, UCHAR *a_output) 
{ 
    UCHAR  *input; 
    UCHAR  *output; 

    input = &(a_input[threadIdx.x * DOCU_SIZE]); 
    output = &(a_output[threadIdx.x * DOCU_SIZE]); 

    for (int i = 0 ; i < 30 ; i++) { 
     if ((input[i] >= 'a') && (input[i] <= 'z')) { 
      output[i] = input[i] - 'a' + 'A'; 
     } 
     else { 
      output[i] = input[i]; 
     } 
    } 
} 


int main(int argc, char** argv) 
{ 
    struct cudaDeviceProp xCUDEV; 

    cudaGetDeviceProperties(&xCUDEV, 0); 


    // Prepare Source 
    memset(pH_TXT, 0x00, DOCU_SIZE * TOTAL); 

    for (int i = 0 ; i < TOTAL ; i++) { 
     strcpy((char*)pH_TXT + (i * DOCU_SIZE), "hello world, i need an apple."); 
    } 

    // Allocate vectors in device memory 
    cudaMalloc((void**)&pD_TXT, DOCU_SIZE * TOTAL); 
    cudaMalloc((void**)&pD_ENC, DOCU_SIZE * TOTAL); 

    // Copy vectors from host memory to device memory 
    cudaMemcpy(pD_TXT, pH_TXT, DOCU_SIZE * TOTAL, cudaMemcpyHostToDevice); 

    // Invoke kernel 
    int threadsPerBlock = BLOCK_SIZE; 
    int blocksPerGrid = (TOTAL + threadsPerBlock - 1)/threadsPerBlock; 

    printf("Total Task is %d\n", TOTAL); 
    printf("block size is %d\n", threadsPerBlock); 
    printf("repeat cnt is %d\n", blocksPerGrid); 

    TEST_Encode<<<blocksPerGrid, threadsPerBlock>>>(pD_TXT, pD_ENC); 

    cudaMemcpy(pH_ENC, pD_ENC, DOCU_SIZE * TOTAL, cudaMemcpyDeviceToHost); 

    // Free device memory 
    if (pD_TXT)   cudaFree(pD_TXT); 
    if (pD_ENC)   cudaFree(pD_ENC); 

    cudaDeviceReset(); 
} 

И когда я изменить значение BLOCK_SIZE от 2 до 1000, я получил следующий раз длительности (от NVIDIA Визуального Profiler)

TOTAL  BLOCKS  BLOCK_SIZE Duration(ms) 
100000  50000  2   28.22 
100000  10000  10   22.223 
100000  2000  50   12.3 
100000  1000  100   9.624 
100000  500   200   10.755 
100000  250   400   29.824 
100000  200   500   39.67 
100000  100   1000  81.268 

Моей видеокарты GeForce GT520 и максимального threadsPerBlock значение 1024, так Я предсказал, что я получаю лучшую производительность, когда BLOCK составляет 1000, но приведенная выше таблица показывает разные результаты.

Я не могу понять, почему Duration не является линейным и как я могу исправить эту проблему. (Или как я могу найти оптимизированное значение Block (Mimimum время Duration)

Answer 1

кажется 2, 10, 50 потоков не использует возможности графического процессора, так как его дизайн, чтобы начать гораздо больше нитей.

Ваш карта имеет вычислительную способность 2.1.

• Максимальное количество потоков резидентов в многопроцессорной = 1536
• Максимальное количество потоков на блок = 1024
• Максимальное количество блоков резидентов в многопроцессорной = 8
• размер Деформация = 32

Есть два вопроса:

1.

Вы пытаетесь занять столько памяти регистров для каждого потока, что он определенно будет аутсорсингом замедлять местный если размеры вашего блока увеличиваются.

2.

Выполните тесты с несколькими из 32, поскольку это размера основы вашей карты и много операций памятей оптимизированы для размеров резьбы с кратным размером основы.

Таким образом, если вы используете только около 1024 (1000 в вашем случае) потоков на блок, 33% вашего gpu простаивает, так как только один блок может быть назначен для SM.

Что произойдет, если вы используете следующие 100% заполняемости?

• 128 = 12 блоков -> поскольку только 8 может быть резидентом в см исполнение блок сериализованные блоки
• 192 = 8 резидентов в см
• 256 = 6 блоков резидентов в см
• 512 = 3 резидентных блока на см