Вычисление Numba Cuda, по-видимому, выполняется медленнее, чем последовательный запуск. Сделал ли я очевидные ошибки?

#performance #cuda #numba

Вопрос:

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

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

https://github.com/ContinuumIO/gtc2017-numba/blob/master/4 — Writing CUDA Kernels.ipynb

который показывает, как параллельно выполнять добавление массивов. Конфигурация системы, которую они использовали для оценки времени, не указана. Для репликации кода я использую Geforce GTX 1080 Ti и процессор Intel Core i7 8700K.

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

 from numba import cuda
import numpy as np
import time
import math

@cuda.jit
def addition_kernel(x, y, out):

    tx = cuda.threadIdx.x
    ty = cuda.blockIdx.x
    block_size = cuda.blockDim.x  
    grid_size = cuda.gridDim.x  

    start = tx  ty * block_size
    stride = block_size * grid_size
    for i in range(start, x.shape[0], stride):
        out[i] = y[i]   x[i]

def add(n, x, y):
    for i in range(n):
        y[i] = y[i]   x[i]


if __name__ =="__main__":
    print(cuda.gpus[0])
    print("")
    n = 100000
    x = np.arange(n).astype(np.float32)
    y = 2 * x
    out = np.empty_like(x)
    x_device = cuda.to_device(x)
    y_device = cuda.to_device(y)
    out_device = cuda.device_array_like(x)



    # Set the number of threads in a block

    threadsperblock = 128

    # Calculate the number of thread blocks in the grid
    blockspergrid = 30#math.ceil(n[0] / threadsperblock)
    # Now start the kernel
    start = time.process_time()
    cuda.synchronize()
    addition_kernel[blockspergrid, threadsperblock](x_device, y_device, out_device)
    cuda.synchronize()
    end = time.process_time()
    out_global_mem = out_device.copy_to_host()
    print("parallel time: ", end - start)

    start = time.process_time()
    add(n,x,y)
    end = time.process_time()
    print("sequential time: ", end-start)
 

Параллельное время составляет в среднем около 0,14 секунды, в то время как код без ядра GPU занимает всего 0,02 секунды.

Мне это кажется довольно странным. Есть ли что-нибудь, что я сделал не так? Или эта проблема не является хорошим примером параллелизма? (что я не думаю, так как вы можете запускать цикл for параллельно)

Что странно, так это то, что я почти не замечаю разницы, если не использую to_device() функции. Насколько я понял, это должно быть важно, так как они избегают связи между процессором и графическим процессором после каждой итерации.

Ответ №1:

addition_kernel компилируется во время выполнения, когда он вызывается в первый раз, так что в середине вашего измеренного времени! Компиляция ядра-довольно трудоемкая операция. Вы можете ускорить вычисление заранее, предоставив типы в Numba.

Обратите внимание, что массивы слишком малы, поэтому вы можете увидеть значительное улучшение графических процессоров. Более того, сравнение с версией процессора на самом деле несправедливо: вы также должны использовать Numba для реализации процессора или, по крайней мере, Numpy (но не интерпретируемый цикл pure-CPython).

Вот пример:

 import numba as nb

@cuda.jit('void(float32[::1], float32[::1], float32[::1])')
def addition_kernel(x, y, out):

    tx = cuda.threadIdx.x
    ty = cuda.blockIdx.x
    block_size = cuda.blockDim.x  
    grid_size = cuda.gridDim.x  

    start = tx  ty * block_size
    stride = block_size * grid_size
    for i in range(start, x.shape[0], stride):
        out[i] = y[i]   x[i]

@nb.njit('void(int64, float32[::1], float32[::1])')
def add(n, x, y):
    for i in range(n):
        y[i] = y[i]   x[i]
 

Комментарии:

1. У вас есть пример того, как я могу предоставить типы данных в numba? После поиска в Google я прочитал, что AOT-компиляция недоступна для cuda numba, но, возможно, вы имеете в виду что-то другое? Что касается справедливости, я согласен, но код, который я добавил для последовательного варианта, был самым медленным методом, который я мог придумать. Поэтому я был особенно удивлен, когда процессор оказался быстрее графического процессора. Однако я заметил, что графический процессор обгоняет процессор после увеличения размера массива (в 10 раз).

2. Я добавил пример. На моей машине это исправляет проблему с кодом. Обратите внимание, что я должен умножить размер (n) на 1000, чтобы тайминги не округлялись до 0. Реализация GPU на моем Nvidia 1660 Super в 2 раза быстрее по сравнению с последовательной реализацией на моем i5-9600KF. Реализация GPU может быть оптимизирована с помощью развертывания, а реализация CPU также может быть оптимизирована с использованием нескольких потоков Numba (см. parallel=True и prange в Numba). Обратите внимание, что два ядра не являются строго эквивалентными, поскольку реализация процессора выполняет вычисления на месте.