параллельная реализация сокращения сумма в OpenCL

Я, проходящей через образец кода NVIDIA предоставляется на link параллельная реализация сокращения сумма в OpenCL

В примере кода ядер (файл oclReduction_kernel.c) reduce4 использует технику

1) разматывания и удаления барьера синхронизации для резьбы id < 32.

2) Кроме этого, код использует проверки blockSize для суммирования данных в локальной памяти. Я думаю, что в OpenCL у нас есть get_local_size(0/1), чтобы узнать размер рабочей группы. Размер блока меня сбивает с толку.

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

источник

2015-07-31 Manish Kumar

Презентация на https://docs.nvidia.com/cuda/samples/6_Advanced/reduction/doc/reduction.PDF относится к CUDA (и, по общему признанию, я считаю, что образцы OpenCL в основном были созданы путем конвертации ядер CUDA в OpenCL как можно скорее), но это может привести к некоторому пониманию, особенно в отношении дальнейших оптимизированных версий reduce5/reduce6. – Marco13

Да, они были только в кудах. Но он не отвечает ни на какие вопросы. –

Спасибо @ Marco13 Я пропустил ур ссылку в ночное время. Может быть, я спал :) –

У вас это в значительной степени объяснено в слайдах 21 и 22 из https://docs.nvidia.com/cuda/samples/6_Advanced/reduction/doc/reduction.pdf, которые @ Marco13 связаны в комментариях.

Как протекает по сокращению, # «активных» потоков уменьшается

Когда с < = 32, то есть только один перекос оставили

Инструкции SIMD-синхронные внутри основы.

Это означает, что при s = < 32:

Мы не должны __syncthreads()

Нам не нужны «если (TID < s)» потому, что он не спасает любую работу

Без разворачивая, все перекосы выполнять каждый его чество из цикла для и если заявление

И https://www.pgroup.com/lit/articles/insider/v2n1a5.htm:

Код фактически выполняется в группах из 32 нитей, то, что NVIDIA называет перекос.

Каждое ядро может выполнять последовательную нить, но ядра выполняют в то, что NVIDIA называет SIMT (Single Instruction, Multiple Thread); все ядра в той же группе выполняют ту же самую команду при том же времени, как и классические процессоры SIMD.

Re 2) blockSize там будет размер рабочей группы.

источник

2015-07-31 23:34:42 doqtor

Когда мы говорим, что код SIMD-4 скомпилирован для GPU. Правильно ли говорить, что :: Будет четыре ядра SIMD, которые будут выполнять ту же инструкцию, и инструкция может быть похожа на float/float4/float16 ?? –

параллельная реализация сокращения сумма в OpenCL

ответ

Смежные вопросы