Понимание слияния в OpenCL

Question

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

Однако, когда я посмотрел на реальный пример, я понятия не имел, почему следующая функция имеет плохую коалесценцию:

__kernel void mapSin(
    __global float* input,
    __global float* input,
    const unsigned int cout)
{
    int x = get_global_id(0);
    int y = get_global_id(1);
    output[x * count + y] = sin(input[x * count + y]);
}

Я предполагаю, что умножение x * count вызывает дискретный доступ к памяти. Я прав? И как улучшить эту функцию, чтобы она обеспечивала лучшее слияние?

Answer 1

Да, вы правы, потому что x * count заставляет каждый рабочий элемент читать память, которая не смежна с другими соседними рабочими элементами. Широкие чтения из памяти используются частично (а остальные отбрасываются, и вы получаете сериализованные чтения для каждого рабочего элемента). x - это «быстрее движущийся» global_id в вашем коде, поэтому попробуйте вместо этого сделать его id (1) и сделать y id (0) (а затем также поменять местами ширину и высоту, где вы ставите ядро ​​в очередь). Это сделает y идентификатором, который изменится на 1 для смежных рабочих элементов, что также сделает ваши обращения к памяти смежными, что является способом получить объединенный доступ к памяти (одно широкое чтение может удовлетворить несколько рабочих элементов; т.е. параллельное чтение) .