Некоторые свойства, названные с помощью `_sync()`, добавлены в CUDA 9; семантика такая же?

Question

В CUDA 9 nVIDIA, похоже, имеет это новое понятие "кооперативные группы"; и по какой-то причине мне не совсем понятно, __ballot() сейчас (= CUDA 9) устарела в пользу __ballot_sync(), Это псевдоним или семантика изменилась?

... похожий вопрос для других встроенных __sync() добавил к своим именам.

Answer 1

Нет, семантика не одинакова. Сами вызовы функций различны, один не является псевдонимом для другого, открыта новая функциональность, и поведение реализации теперь отличается в архитектуре Volta и предыдущих архитектурах.

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

Большинство встроенных функций деформации работают, предоставляя ожидаемые результаты только для потоков, которые фактически участвуют (то есть фактически активны для выполнения этой инструкции в этом цикле). Теперь программист может четко определить, какие потоки должны участвовать, с помощью нового mask параметр. Однако есть некоторые требования, в частности, к Pascal и предыдущим архитектурам. Из руководства по программированию:

Обратите внимание, однако, что для Pascal и более ранних архитектур все потоки в mask должна выполнить ту же самую внутреннюю инструкцию деформации при конвергенции, и объединение всех значений в маске должно быть равно активной маске деформации.

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

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