Блог

Я создаю ядро для поэлементного умножения двух матриц, но, по крайней мере, с моими конфигурациями, мое ядро OpenCL работает быстрее, только когда размер каждой матрицы превышает 2 ГБ. Итак, мне было интересно, связано ли это с моим наивным ядром (см. Ниже) или из-за природы поэлементных операций, означающих, что поэлементные операции не выигрывают от использования графических процессоров.

Спасибо за ваш вклад!

ядро:

 KERNEL_CODE = """
// elementwise multiplication: C = A .* B.
__kernel void matrixMul(
        __global float* C,
        __global float* A,
        __global float* B,
        int width, int height)
{
    // ID
    int x = get_global_id(0);
    int y = get_global_id(1);

    // Multiplying
    C[y * height   x ] = A[y * height   x] * B[y * height   x];
}
"""
  

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

Такого рода операции имеют N провалов, но 3N транзакций с памятью, поэтому они будут полностью ограничены пропускной способностью памяти. Нет возможности для повторного использования данных, поэтому верхняя граница ускорения по сравнению с эталонной версией CPU — это отношение пропускной способности GPU к CPU. Это число редко превышает 10 раз и может довольно быстро уменьшиться из-за затрат на перемещение данных в память графического процессора и из нее. Вообще говоря, такого рода операции лучше всего «объединять» с другими O (N) операциями для повышения производительности. Обычно вы бы никогда не стали просто вычислять произведение Адамара в одном ядре, вы бы сделали это как часть серии из O (N) операций в одном ядре. Итак, нет, это не лучший кандидат для ускорения, даже если бы ядро было оптимальным.

И ваше ядро определенно таковым не является. Вы выполняете 3 операции ввода-вывода за каждый ФЛОП, что является огромным штрафом. Вы определенно могли бы что-то сделать, чтобы улучшить это, но что именно, будет полностью зависеть от того, на каком оборудовании это будет выполняться.

Говоря о поэлементных операциях: это зависит от устройства. Например, графические процессоры NVidia используют скалярные процессоры (со скалярными инструкциями), векторизация не требуется. Напротив, ATI использует 5d (или 4d) VLIW-процессоры, и для них решающее значение имеет векторизация. Однако иногда это может выполняться компилятором вместо использования векторных типов данных непосредственно в коде, но это первое, что нужно сделать при оптимизации для графических процессоров ATI.

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

Опубликованное вами ядро должно быть по крайней мере таким же быстрым, как ядро процессора. Но вы вообще не используете объединенные обращения к памяти!

Это снижает вашу производительность.

Однако, как заявил @talonmies. Это не лучший вариант для графического процессора. Вы теряете все свое время на копирование в память.