Позвольте мне начать с указания на то, что разделяемая память - это, прежде всего, абстракция модели программирования, с помощью которой раскрывается определенная функция аппаратного обеспечения (быстрая память на кристалле). В модели программирования CUDA каждый блок в сетке (запуск ядра) получает тот же объем разделяемой памяти. Сколько это зависит, зависит от количества статически распределенной разделяемой памяти, необходимой для функции ядра, а также от любой дополнительной динамической разделяемой памяти, указанной при запуске ядра.
- имеет ли каждый SM одинаковый объем общей памяти в одном и том же графическом процессоре?
Да, это так. Тем не менее, это не так важно для программирования CUDA, как вы думаете, потому что:
- Как SM распределяет разделяемую память между блоками? Распределяется ли он одинаково (например, если есть 2 блока, то каждый блок получает половину общей памяти внутри SM независимо от того, сколько фактически используется), или это основано на потребностях?
Когда вы запускаете ядро, вы указываете, сколько общей памяти требуется каждому блоку. Затем он сообщает, сколько блоков может поместиться на каждом мультипроцессоре. Таким образом, дело не в том, что количество блоков определяет количество разделяемой памяти, которое получает каждый блок, а в другом: количество разделяемой памяти, необходимое для каждого блока, является одним из факторов, определяющих, сколько блоков может находиться на каждом многопроцессорном компьютере.
Вы, возможно, захотите прочитать о скрытии задержек и загруженности, поскольку это довольно фундаментальные темы, когда речь идет о программировании на GPU. Для получения дополнительной информации о подсистемах памяти различных архитектур графических процессоров см. Руководство по программированию CUDA .
- Мое понимание банка общей памяти таково: общая память делится на 32 одинаково больших банка памяти. Так значит ли это для каждого блока (т. Е. У каждого блока есть свои 32 банка) или для каждого SM?
В конце концов, из-за SIMD (SIMT) природы ядер GPU, фактическое выполнение программы происходит в перекосах. Когда такая деформация (в настоящее время это фактически означает группу из 32 потоков) выполняет доступ к общей памяти, конфликты банков будут проблемой, поскольку обслуживается запрос общей памяти, сгенерированный этой инструкцией. На самом деле не задокументировано, могут ли запросы к разделяемой памяти обслуживаться несколькими искажениями параллельно. Я полагаю, что для каждого запроса SM существует только один модуль, и поэтому ответ - нет.
- Если я выполняю CudaMemcpy из / в общую память более чем из одного слова, считается ли это как одна транзакция или несколько транзакций? И может ли это вызвать банковские конфликты?
Вы не можете cudaMemcpy() в общую память. Общая память доступна только для потоков устройств одного и того же блока и сохраняется только до тех пор, пока работает этот блок.