Локальная, глобальная, постоянная и разделяемая память

Question

Я прочитал некоторую документацию CUDA, которая относится к локальной памяти. (В основном это ранняя документация.) Свойства устройства сообщают о размере локальной памяти (для каждого потока). Что означает "локальная" память? Что такое "локальная" память? Где находится "локальная" память? Как я могу получить доступ к "локальным" мемам? это __device__ память нет?

Свойства устройства также сообщают: глобальный, общий и постоянный размер памяти. Верны ли эти утверждения:Глобальная память __device__ объем памяти. Он имеет область видимости сетки и время жизни сетки (ядра).Постоянная память есть __device__ __constant__ объем памяти. Он имеет область видимости сетки и время жизни сетки (ядра). Общая память __device__ __shared__ объем памяти. Он имеет единичную область видимости и время жизни этого блока (потоков).

Я думаю, что общая память - это память SM. то есть память, к которой имел доступ только тот единственный SM. Ресурс, который довольно ограничен. Разве SM не назначен кучу блоков одновременно? Означает ли это, что SM может чередовать выполнение разных блоков (или нет)? т.е. запускать блоки *A*, пока они не остановятся. Затем запускайте блоки *B*, пока они не остановятся. Затем вернитесь, чтобы снова заблокировать потоки *A*. ИЛИ Запускает ли SM набор потоков для блока *A*, пока они не остановятся. Затем заменяется другой набор потоков блока *A*. Этот обмен продолжается до тех пор, пока блок *A* не будет исчерпан. Тогда и только тогда начинается работа на блоке *B*. Я спрашиваю из-за общей памяти. Если один SM меняет код в 2 разных блоках, то как SM быстро меняет местами блоки памяти? (Я думаю, что более поздняя версия senerio верна, и не происходит подкачки в / из разделяемой памяти. Блок *A* выполняется до завершения, затем блок *B* начинает выполнение. Примечание: блок *A* может быть другим ядро, чем блок *B*.)

Answer 1

Из раздела 5.3.2.2 Руководства по программированию в CUDA C мы видим, что локальная память используется в нескольких случаях:

• Когда у каждого потока есть несколько массивов, но их размер неизвестен во время компиляции (поэтому они могут не помещаться в регистры)
• Когда размер массивов известен во время компиляции, и этот размер слишком велик для памяти регистров (это также может случиться с большими структурами)
• Когда ядро ​​уже израсходовало всю память регистров (поэтому, если мы заполнили регистры nintс n+1го int войдет в локальную память) - это последний случай - разлитие регистров, и этого следует избегать, потому что:

"Локальная" память фактически живет в глобальном пространстве памяти, что означает, что чтение и запись в нее сравнительно медленны по сравнению с регистровой и разделяемой памятью. Вы будете получать доступ к локальной памяти каждый раз, когда будете использовать переменную, массив и т. Д. В ядре, которое не помещается в регистры, не является общей памятью и не передается в качестве глобальной памяти. Вам не нужно ничего делать явным образом, чтобы использовать его - на самом деле вы должны попытаться свести к минимуму его использование, так как регистры и разделяемая память намного быстрее.

Редактировать: Re: общая память, вы не можете иметь два блока, которые обмениваются общей памятью или смотрят на общую память друг друга. Поскольку порядок выполнения блоков не гарантирован, если вы попытаетесь это сделать, вы можете связать SMP на несколько часов, ожидая выполнения другого блока. Точно так же два ядра, работающие на устройстве одновременно, не могут видеть память друг друга, если только это не глобальная память, и даже тогда вы играете с огнем (условий гонки). Насколько я знаю, блоки / ядра не могут действительно посылать друг другу "сообщения". Ваш сценарий на самом деле не имеет смысла, поскольку порядок выполнения для блоков будет каждый раз отличаться, и плохая практика - блокировать блок, ожидая другого.