Как интерпретировать эти результаты для среднего фильтра для серийных версий GPU и CPU?

Question

Я реализовал изображение Mean Filter код для CPU serial версии и NVIDIA GPU parallel версии. У меня есть время выполнения (см. результаты тестовых случаев и спецификации устройств. Почему case 2 имеет ускорение highest, а case 3 имеет lowest ускорение?

Конфигурация исполнения графического процессора

        int block_size = 32;
        int grid_size = width/block_size; //width of the image in pixels
        dim3 dimBlock(block_size, block_size, 1);
        dim3 dimGrid(grid_size, grid_size, 1);

измерение времени для кода GPU

        clock_t start_d=clock();
        meanFilter_d <<< dimGrid, dimBlock >>> (image_data_d, result_image_data_d, width, height, half_window);
        cudaThreadSynchronize();
        clock_d end_d=clock();

измерение времени для кода ЦП (однопоточный)

        clock_t start_h = clock();
        meanFilter_h(data, result_image_data_h1, width, height, window_size);
        clock_t end_h = clock();

код хоста

void meanFilter_h(unsigned char* raw_image_matrix,unsigned char* filtered_image_data,int image_width, int image_height, int window_size)
{
    // int size = 3 * image_width * image_height;
    int half_window = (window_size-window_size % 2)/2;
    for(int i = 0; i < image_height; i += 1){
        for(int j = 0; j < image_width; j += 1){
            int k = 3*(i*image_height+j);
            int top, bottom, left, right; 
            if(i-half_window >= 0){top = i-half_window;}else{top = 0;}// top limit
            if(i+half_window <= image_height-1){bottom = i+half_window;}else{bottom = image_height-1;}// bottom limit
            if(j-half_window >= 0){left = j-half_window;}else{left = 0;}// left limit
            if(j+half_window <= image_width-1){right = j+half_window;}else{right = image_width-1;}// right limit
            double first_byte = 0; 
            double second_byte = 0; 
            double third_byte = 0; 
            // move inside the window
            for(int x = top; x <= bottom; x++){
                for(int y = left; y <= right; y++){
                    int pos = 3*(x*image_height + y); // three bytes
                    first_byte += raw_image_matrix[pos];
                    second_byte += raw_image_matrix[pos+1];
                    third_byte += raw_image_matrix[pos+2];
                }
            }
            int effective_window_size = (bottom-top+1)*(right-left+1);
            filtered_image_data[k] = first_byte/effective_window_size;
            filtered_image_data[k+1] = second_byte/effective_window_size;
            filtered_image_data[k+2] =third_byte/effective_window_size;


        }
    }
}

код устройства

__global__ void meanFilter_d(unsigned char* raw_image_matrix, unsigned char* filtered_image_data, int image_width, int image_height, int half_window)
{
    int j = blockIdx.x * blockDim.x + threadIdx.x;
    int i = blockIdx.y * blockDim.y + threadIdx.y;

    if (i < image_height && j < image_width){
        int k = 3*(i*image_height+j);
        int top, bottom, left, right; 
        if(i-half_window >= 0){top = i-half_window;}else{top = 0;}// top limit
        if(i+half_window <= image_height-1){bottom = i+half_window;}else{bottom = image_height-1;}// bottom limit
        if(j-half_window >= 0){left = j-half_window;}else{left = 0;}// left limit
        if(j+half_window <= image_width-1){right = j+half_window;}else{right = image_width-1;}// right limit
        double first_byte = 0; 
        double second_byte = 0; 
        double third_byte = 0; 
        // move inside the window
        for(int x = top; x <= bottom; x++){
            for(int y = left; y <= right; y++){
                int pos = 3*(x*image_height + y); // three bytes
                first_byte += raw_image_matrix[pos];
                second_byte += raw_image_matrix[pos+1];
                third_byte += raw_image_matrix[pos+2];
            }
        }
        int effective_window_size = (bottom-top+1)*(right-left+1);
        filtered_image_data[k] = first_byte/effective_window_size;
        filtered_image_data[k+1] = second_byte/effective_window_size;
        filtered_image_data[k+2] =third_byte/effective_window_size;
    }
}

Видно, что оба размера изображения с ядром 3×3 медленнее, чем 5*5 ядро. Случай 1 имеет больше параллелизма, чем случай 3, из-за большего размера изображения. Таким образом, использование устройства в случае 1 выше, чем в случае 3. Но я понятия не имею, интерпретировать дальше. Пожалуйста, дайте мне некоторые идеи.

Answer 1

Первое, на что следует обратить внимание: что вы измеряете и самое важное как ? Из твоего вопроса невозможно особо понять, как.

Во всяком случае, я настоятельно рекомендую вам взглянуть на это, это очень простая и полезная статья Марка Харриса, в которой объясняются некоторые полезные практики для выборки времени выполнения кода на стороне устройства (например, передачи памяти CUDA). , ядра и т. д.).

Кстати, попытка получить ускорение CPU / GPU - довольно сложная тема, это связано с действительно разной природой двух архитектур. Даже если ваши коды CPU и GPU, по-видимому, делают одно и то же, существует множество факторы, которые, возможно, вы хотите принять во внимание (например, ядра процессора, потоковые мультипроцессоры GPU и ядра на SM). Здесь Роберт Кровелла дает отличный ответ на подобную проблему, как он говорит:

Если вы делаете какие-либо заявления о том, что «графический процессор быстрее, чем центральный процессор к XX», то, IMO, вам рекомендуется сравнивать только те коды, которые выполняют ту же работу, и эффективно и результативно использовать базовые архитектуры (как для ЦП, так и для ЦП). GPU). Например, в случае с процессором вы, безусловно, должны использовать многопоточный код, чтобы использовать преимущества нескольких ядер CPU, которые предлагает большинство современных процессоров. Подобные претензии, скорее всего, будут восприняты со скептицизмом, так что, вероятно, лучше избегать их, если это не суть вашего намерения.

Я предлагаю вам взглянуть на это обсуждение тоже.

После некоторых посылок я не думаю, что вы можете считать эти ускорения надежными (на самом деле они кажутся мне немного странными).
Пытаясь истолковать то, что вы пытались сказать:

Видно, что оба размера изображения с ядром 3 × 3 медленнее

Может быть, вы хотели сказать, что в 3х3 вы получили меньшее ускорение w.r.t. те для размера окна 5x5. Постарайтесь быть более точным.

Почему случай 2 имеет самое высокое ускорение, а случай 3 имеет самое низкое ускорение?

Ну, на самом деле трудно что-то сделать на основании предоставленной вами недостоверной информации.

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

---

РЕДАКТИРОВАТЬ:

Ну, я думаю, вы должны принять меры более точно.

• Сначала я бы порекомендовал вам использовать более точную альтернативу clock(). Посмотрите ответы здесь и ссылку на C ++, я предлагаю вам рассмотреть вопрос об использовании

std::chrono::system_clock::now()

std::chrono::high_resolution_clock::now();

• Тогда я повторю вас, чтобы прочитать статью (ссылка выше) Марка Харриса. Здесь он говорит

  Проблема с использованием точек синхронизации хост-устройства, таких как cudaDeviceSynchronize(), заключается в том, что они блокируют конвейер графического процессора. По этой причине CUDA предлагает относительно легкую альтернативу таймерам ЦП через API событий CUDA. API событий CUDA включает вызовы для создания и уничтожения событий, записи событий и вычисления истекшего времени в миллисекундах между двумя записанными событиями.

Это означает, что фактические результаты по предоставленным мерам могут быть немного "искажены" при использовании cudaDeviceSynchronize(). Кроме того, нет необходимости использовать механизм синхронизации, если вы используете простой cudaMemcpy, поскольку это синхронный вызов.

• Также подумайте о включении передач H2D / D2H, по моему мнению, важно учитывать эти издержки при сравнении CPU / GPU (но этот выбор остается за вами);
• О мерах, которые вы дали на картинке, являются ли они прямым результатом или среднее число повторных различных казней (возможно, отбрасывая значения расходов)?

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

Кстати, вы сказали

Случай 1 имеет больший параллелизм, чем случай 3, из-за большего размера изображения.Поэтому использование устройства для случая 1 выше, чем для случая 3.

Я не согласен с этим, так как вы int grid_size = width/block_size;

Случай 1: grid_size = 640/32 = 20

CASE 2: grid_size = 1280/32 = 40

Так что у вас больше параллелизма вСлучай 2. Но так как у вас всего 2 SM, это может быть причиной того, что время может быть выше, чем вы ожидали.Другими словами, у вас есть больше блоков (40 * 40), ожидающих вычисления двух SM.