Почему скорость копирования памяти CUDA ведет себя так, что некоторые постоянные издержки драйвера?

Question

При работе с памятью в CUDA на моем старом GeForce 8800GT всегда возникают странные издержки в 0, 04 мс. Мне нужно передать ~1-2K в постоянную память моего устройства, работать с этими данными на нем и получить только одно значение с плавающей запятой от устройства.

У меня есть типичный код с использованием вычисления GPU:

//allocate all the needed memory: pinned, device global
for(int i = 0; i < 1000; i++)
{
    //Do some heavy cpu logic (~0.005 ms long)        
    cudaMemcpyToSymbolAsync(const_dev_mem, pinned_host_mem, mem_size, 0, cudaMemcpyHostToDevice);
    my_kernel<<<128, 128>>>(output);
    //several other calls of different kernels
    cudaMemcpy((void*)&host_output, output, sizeof(FLOAT_T), cudaMemcpyDeviceToHost);
    // Do some logic with returned value 
}

Я решил измерить скорость работы с памятью GPU этим кодом (прокомментировал все вызовы ядра, добавил cudaDeviceSynchronize вызов):

//allocate all the needed memory: pinned, device global
for(int i = 0; i < 1000; i++)
{
    //Do some heavy cpu logic (~0.001 ms long)        
    cudaMemcpyToSymbolAsync(const_dev_mem, pinned_host_mem, mem_size, 0, cudaMemcpyHostToDevice);
    cudaMemcpyAsync((void*)&host_output, output, sizeof(FLOAT_T), cudaMemcpyDeviceToHost);
    cudaDeviceSynchronize();
    // Do some logic with returned value 
}

Я измерил время выполнения цикла и получил ~0, 05 с (то есть 0, 05 мс за итерацию). Странно то, что когда я пытаюсь сделать больше работы с памятью (добавляя дополнительные вызовы cudaMemcpyToSymbolAsync и cudaMemcpyAsync), я получаю дополнительное время <0, 01 мс на вызов. Это соответствует исследованию этого парня: http://www.cs.virginia.edu/~mwb7w/cuda_support/memory_transfer_overhead.html

Он также получил эти 0, 01 мс за передачу блока 1К на графический процессор. Так откуда взялись эти 0, 04 мс (0, 05 - 0, 01) служебные данные? Есть идеи? Может быть, я должен попробовать этот код на новой карте?

Мне кажется, что после cudaDeviceSynchronize и кода процессора моя GeForce переходит в какой-то режим энергосбережения или что-то в этом роде.

Answer 1

Я рекомендую вам увеличить количество потоков, которые вы реализуете

    //Use malloc() to allocate memory on CPU. 
    //Change mem_size to the total memory to be tranferred to GPU.        
    cudaMemcpyToSymbolAsync(const_dev_mem, pinned_host_mem, mem_size, 0, cudaMemcpyHostToDevice);
    dim3 dimBlock(128,2);
    dim3 dimGrid(64000,1);
    my_kernel<<<dimGrid, dimBlock>>>(output);
    //several other calls of different kernels
    //change size field to 1000*sizeof(FLOAT_T)
    cudaMemcpy((void*)&host_output, output, sizeof(FLOAT_T), cudaMemcpyDeviceToHost);
    // Do some logic with returned value 

Если код дает сбой (из-за большего количества потоков или большего количества памяти GPU), используйте циклы. Но сделай их меньше.