Почему рабочие группы OpenCL 3-мерные?

Question

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

For example, this person discovered that switching axis sped up his execution with a factor of two.

Одна из гипотез, у меня есть то, что OpenCL хочет тривиальные отношения между идентификатором рабочего элемента и поиском памяти, чтобы обеспечить предсказуемые операции с памятью, которые могут быть I/O оптимизированы.

Answer 1

Рабочие группы не обязательно должны быть трехмерными, если ваше приложение/алгоритм не требует этого. Вы можете указать 1, 2 или 3 измерения - и, несомненно, больше в будущем. Поэтому используйте меньшие размеры, когда, естественно, подходит для вашего приложения.

Итак, почему спецификация позволит увеличить размеры? Как вы указали, более высокие размеры можно эмулировать с использованием одного измерения. Одним из примеров может быть трехмерное моделирование N-Body для физико-молекулярного моделирования.

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

Что касается увеличения производительности 2x в вашем примере: это повышение было результатом гораздо лучшего шаблона доступа к памяти, а не аппаратного обеспечения, которое по своей природе было ужасным при работе в 2D рабочей группе. Ответ на этот вопрос объясняет способы дальнейшей оптимизации ядра, которые являются отличными стратегиями для сегодняшнего оборудования gpu.

Более тонкое преимущество использования трехмерных рабочих групп заключается в том, что будущему оборудованию может не понадобиться эмулировать дополнительные размеры. Возможно, память, процессор и т. Д. Будут адаптированы к трехмерным рабочим группам и уменьшат или устраняют штраф за неудачные шаблоны доступа к памяти. Если вы пишете свой код с помощью 1D-групп, вы упустите потенциальное повышение производительности на этих платформах. Даже сегодня можно создавать чипы FPGA/ASIC для работы с 3D рабочими группами лучше, чем графические процессоры.

Answer 2

Что действительно говорит вам, что разрешены только 3 измерения?

clEnqueueNDRangeKernel() использует целое число без знака, чтобы указать количество измерений, и использует массив целых чисел без знака для каждого размера измерения.

В спецификации OpenCL указано, что максимальное число измерений составляет . Реализация определена как константа CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, которая на практике часто бывает 3, но может быть что угодно. Это просто вопрос удобства, так как большинство вычислительных проблем работают с данными «реального мира», которые имеют от 1 до 3 измерений.

Кроме того, никто не заставляет вас использовать 3. Большинство приложений используют 1 и 2, и работают отлично.

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

Вы можете эмулировать его с умными преобразованиями индексов, но используя его в качестве API, он намного проще и читабельнее.