Почему мой инклюзивный код сканирования в 2 раза быстрее на ЦП, чем на ГП?

Я написал короткую программу CUDA, в которой используется высокооптимизированная библиотека CUB, чтобы продемонстрировать, что одно ядро ​​из старого, Четырехъядерный процессор Intel Q6600 (все четыре предположительно способны выполнять ~30 GFLOPS/сек) может выполнять инклюзивное сканирование (или кумулятивную/префиксную сумму, если хотите) на 100 000 элементов быстрее, чем Nvidia 750 Ti (предположительно способный к 1306 GFLOPS/сек). сек одинарной точности). Почему это так?

Исходный код:

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <cub/cub.cuh>

#include <stdio.h>
#include <time.h>
#include <algorithm>


#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort = true)
{
    if (code != cudaSuccess)
    {
        fprintf(stderr, "GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
        if (abort) exit(code);
    }
}

void fillArrayWithRandom(float* inputArray, int inputN)
{
    for (int i = 0; i < inputN; i++)
    {
        inputArray[i] = (float)rand() / float(RAND_MAX);
    }
}

void inclusiveSum_CPU(float *inputArray, float *inputSummedArray, int inputN)
{
    for (int i = 0; i < inputN; i++)
    {
        if (i > 0)
        {
            inputSummedArray[i] = inputSummedArray[i - 1] + inputArray[i];
        }
        else
        {
            inputSummedArray[i] = inputArray[i];
        }
    }
}

int main()
{
    int N = 100000; //1 hundred thousand elements
    float numSimulations = 10000;

    //Make Host Arrays
    float* testArray_CPU = (float *)malloc(sizeof(float)*N);
    fillArrayWithRandom(testArray_CPU, N);
    float* testArrayOutput_CPU = (float *)malloc(sizeof(float)*N);

    //Make GPU Arrays
    float* testArray_GPU;
    gpuErrchk(cudaMalloc(&testArray_GPU, N*sizeof(float)));
    gpuErrchk(cudaMemcpy(testArray_GPU, testArray_CPU, N*sizeof(float), cudaMemcpyHostToDevice));
    float* testArrayOutput_GPU;
    gpuErrchk(cudaMalloc(&testArrayOutput_GPU, N*sizeof(float)));

    //Initiate the benchmark variables
    clock_t begin_CPU, end_CPU;
    float time_spent_GPU, time_spent_CPU;

    //GPU prep
    void     *d_temp_storage = NULL;
    size_t   temp_storage_bytes = 0;
    cub::DeviceScan::InclusiveSum(d_temp_storage, temp_storage_bytes, testArray_GPU, testArrayOutput_GPU, N);
    gpuErrchk(cudaMalloc(&d_temp_storage, temp_storage_bytes));

    //GPU Timing

    cudaEvent_t start, stop;
    gpuErrchk(cudaEventCreate(&start));
    gpuErrchk(cudaEventCreate(&stop));
    gpuErrchk(cudaEventRecord(start, 0));
    for (int i = 0; i < numSimulations; i++)
    {
        cub::DeviceScan::InclusiveSum(d_temp_storage, temp_storage_bytes, testArray_GPU, testArrayOutput_GPU, N);
    }
    gpuErrchk(cudaDeviceSynchronize());
    gpuErrchk(cudaEventRecord(stop, 0));
    gpuErrchk(cudaEventSynchronize(stop));
    gpuErrchk(cudaEventElapsedTime(&time_spent_GPU, start, stop));

    cudaError_t error = cudaGetLastError();
    if (error != cudaSuccess)
    {
        printf("CUDA error: %s\n", cudaGetErrorString(error));
        exit(-1);
    }

    time_spent_GPU = (float)(time_spent_GPU / 1000);
    float avg_GPU = time_spent_GPU / numSimulations;
    printf("Avg. GPU Simulation Time: %.17g [sim/sec]\n", avg_GPU);

    //CPU Timing
    begin_CPU = clock();
    for (int i = 0; i < numSimulations; i++)
    {
        inclusiveSum_CPU(testArray_CPU, testArrayOutput_CPU, N);
    }
    end_CPU = clock();
    time_spent_CPU = (float)(end_CPU - begin_CPU) / CLOCKS_PER_SEC;
    float avg_CPU = time_spent_CPU / numSimulations;
    printf("Avg. CPU Simulation Time: %.17g [sim/sec]\n", avg_CPU);

    printf("GPU/CPU Timing:%.17gx \n", avg_GPU / avg_CPU);

    return 0;
}

И вывод, когда я запускаю его на своей машине:

Сред. Время моделирования графического процессора: 0,0011999999405816197 [сим/сек]

Сред. Время моделирования ЦП: 0,00059999997029080987 [сим/сек]

Время GPU/ЦП: 2x

Кроме того, вот мои флаги компиляции и вывод:

1>------ Build started: Project: speedTest, Configuration: Debug Win32 ------
1>  Compiling CUDA source file kernel.cu...
1>  
1>  C:\Users\Owner\Documents\Visual Studio 2013\Projects\speedTest\speedTest>"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\bin\nvcc.exe" -gencode=arch=compute_50,code=\"sm_50,compute_50\" --use-local-env --cl-version 2013 -ccbin "C:\Program Files (x86)\Microsoft Visual Studio 12.0\VC\bin" -rdc=true -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\include"  -G   --keep-dir Debug -maxrregcount=0  --machine 32 --compile -cudart static  -g   -DWIN32 -D_DEBUG -D_CONSOLE -D_MBCS -Xcompiler "/EHsc /W3 /nologo /Od /Zi /RTC1 /MDd  " -o Debug\kernel.cu.obj "C:\Users\Owner\Documents\Visual Studio 2013\Projects\speedTest\speedTest\kernel.cu" 
1>  kernel.cu
1>  
1>  C:\Users\Owner\Documents\Visual Studio 2013\Projects\speedTest\speedTest>"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\bin\nvcc.exe" -dlink -o Debug\speedTest.device-link.obj -Xcompiler "/EHsc /W3 /nologo /Od /Zi /RTC1 /MDd  " -L"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\lib\Win32" cudart.lib kernel32.lib user32.lib gdi32.lib winspool.lib comdlg32.lib advapi32.lib shell32.lib ole32.lib oleaut32.lib uuid.lib odbc32.lib odbccp32.lib  -gencode=arch=compute_50,code=sm_50 -G --machine 32 Debug\kernel.cu.obj 
1>  cudart.lib
1>  kernel32.lib
1>  user32.lib
1>  gdi32.lib
1>  winspool.lib
1>  comdlg32.lib
1>  advapi32.lib
1>  shell32.lib
1>  ole32.lib
1>  oleaut32.lib
1>  uuid.lib
1>  odbc32.lib
1>  odbccp32.lib
1>  kernel.cu.obj
1>  speedTest.vcxproj -> C:\Users\Owner\Documents\Visual Studio 2013\Projects\speedTest\Debug\speedTest.exe
1>  copy "C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\bin\cudart*.dll" "C:\Users\Owner\Documents\Visual Studio 2013\Projects\speedTest\Debug\"
1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\bin\cudart32_65.dll
1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.5\bin\cudart64_65.dll
1>          2 file(s) copied.
========== Build: 1 succeeded, 0 failed, 0 up-to-date, 0 skipped ==========

person tantrev    schedule 09.12.2014    source источник
comment
Здравствуй, враждебный незнакомец! На самом деле всегда было 100 000 — я просто перепутал комментарий в исходной ссылке, которую разместил, но через несколько минут исправил. И да, точное 2-кратное ускорение было неожиданным, но это действительно то, как печенье рассыпалось. Я чувствую, что функция clock() немного расщепляет волосы, но, тем не менее, я обновил код с подпрограммами синхронизации CUDA. Я пропустил проверку ошибок, потому что смысл примера был в простоте, а не в качественном коде. И в соответствии с просьбой я увеличил количество симуляций (хотя это не оказало существенного влияния) и включил код.   -  person tantrev    schedule 09.12.2014
comment
Результат нового запуска с синхронизацией CUDA: Avg. Время моделирования GPU: 0,0010889752302318811 [sim/sec] Ср. Время моделирования CPU: 0,0005685999640263617 [sim/sec] Время GPU/CPU: 1,9151870012283325x Кроме того, какую карту CUDA вы используете?   -  person tantrev    schedule 09.12.2014
comment
Да, мой плохой, 100 000 элементов. Комментарии на самом деле все щепетильны, за исключением того, что я не могу воспроизвести ваши результаты, поэтому, работая с этими гнидами, возможно, появится какая-то дополнительная подсказка, объясняющая, почему ваши результаты кажутся обратными. (Я предполагаю, что вы думаете, что они могут быть отсталыми, иначе вы, вероятно, не опубликовали бы этот вопрос.) Я почти убежден, что ваш код может продемонстрировать существенное ускорение GPU по сравнению с CPU, просто не уверен, почему это не так в вашем настраивать. Проверка ошибок касается не только производственного кода, но и устранения проблем. Это может пролить свет.   -  person Robert Crovella    schedule 09.12.2014
comment
Абсолютно, я ценю помощь. Я только что добавил проверку на ошибки во всех подпрограммах, которые знаю как (я не совсем уверен, как можно было бы проверять ошибки в подпрограмме CUB), а также проверку на наличие ошибок CUDA и ничего не смог найти. Возможно, ваше ускорение связано с использованием очень мощного графического процессора и/или менее мощного процессора (хотя мой Q6600 на самом деле не такой продвинутый, это может быть проблема AMD и Intel?).   -  person tantrev    schedule 09.12.2014
comment
Я тестировал на Quadro5000. Он может быть примерно в 1,5 раза быстрее, чем GTX750Ti (исходя из характеристик пропускной способности памяти), но, грубо говоря, он все равно должен указывать на то, что возможно. И мой процессор был Xeon W570, поэтому он, вероятно, должен быть быстрее, чем ваш Q6600.   -  person Robert Crovella    schedule 09.12.2014
comment
Интересно, это действительно странно. Может быть, это может быть вариант компилятора? Я только что опубликовал свои текущие флаги и вывод компилятора из Visual Studio 2013.   -  person tantrev    schedule 09.12.2014
comment
Да это оно. Не создавайте отладочную конфигурацию. Создайте конфигурацию выпуска. Переключатель отладки для кода устройства (-G) блокирует большинство оптимизаций компилятора, и это может существенно повлиять на скорость выполнения. В качестве дополнительного комментария окна могут иметь здесь небольшой эффект. Если вы хотите сравнить производительность, вы можете смягчить этот эффект, уменьшив количество итераций до 100, но увеличив количество элементов со 100 000 до 1 000 000. Это уменьшит количество вызовов ядра, но увеличит объем работы на ядро, чтобы амортизировать накладные расходы Windows. Но главное переключатель -G   -  person Robert Crovella    schedule 09.12.2014
comment
Фантастика, спасибо! Я играю с Visual Studio, чтобы попытаться отключить флаг, и сообщу вам о своих результатах. Спасибо еще раз.   -  person tantrev    schedule 09.12.2014
comment
Я бы не стал пытаться идти за флагом напрямую. Просто создайте конфигурацию выпуска вместо конфигурации отладки. Это раскрывающийся список на панели инструментов VS. Это автоматически удалит переключатель -G.   -  person Robert Crovella    schedule 09.12.2014
comment
Я только что обнаружил, что правильно, как вы мне сказали! Ха, я только что получил 20-кратное ускорение от графического процессора. Задача решена! Спасибо. Я обновлю вопрос. Мне нравится, как я превратил эту очень простую задачу в излишне сложную.   -  person tantrev    schedule 09.12.2014
comment
Кто-нибудь, добавьте ответ на этот фарс. Либо так, либо удаляй.   -  person talonmies    schedule 09.12.2014
comment
Ты всегда смешишь меня @talonmies Я рад, что ты не покинул здание.   -  person Robert Crovella    schedule 09.12.2014


Ответы (1)


Благодаря Роберту Кровелле оказалось, что я использовал печально известный медленный режим «Отладка», а не режим «Выпуск».

person tantrev    schedule 09.12.2014
comment
Интересно, что «режим отладки» замедляет работу графического процессора больше, чем замедляет работу процессора. Я не совсем припоминаю, чтобы в документах NVIDIA это упоминалось где-либо. - person einpoklum; 20.04.2015