О константном указателе и передаче параметров в CUDA
Как nvcc обрабатывает константные указатели в ядрах?
Согласно nvidia, добавив const и ограничивать для указателей во время передачи параметров включить NVCC для агрессивной оптимизации, строго ли это соответствует пути C / C ++?
Предполагая, что указатель A указывает на буфер данных, который будет часто обновляться, возможно, другими потоками / потоками, но содержимое не будет изменено во время этого тестового вызова ядра:
test<<<blocks, threads>>>(const int *__restrict__ A, int *__restrict__ B);
Тогда сможет ли NVCC сохранить правильность этого: загружать обновленные данные в A при каждом обращении к ядру, а не загружать некоторые предварительно кэшированные устаревшие данные?
Решение
const работает как C ++. const переменная не может быть изменена, и это проверяется во время компиляции компилятором. Компилятор проверяет правильность констант только для заданной области, поскольку константность может быть изменена с помощью приведения в стиле C.
restrict работает на С способом. Когда вы помечаете указатели как ограничивающие, компилятор предполагает, что эти указатели не являются псевдонимами. Это ваш заданный факт, и компилятор не будет проверять, является ли этот факт истинным или нет.
Возвращаясь к вашему вопросу, NVCC не будет проверять правильность записи и чтения глобальной памяти между запусками ядра. Поскольку запуск ядра в CUDA выполняется асинхронно, вы должны быть уверены, что ядра, модифицирующие эти области памяти, не выполняются одновременно. Вы можете добиться этого с помощью синхронных копий памяти и / или cudaDeviceSynchronize(), Если вы запускаете эти ядра одновременно, то нет никакого способа убедиться, что все изменения, внесенные разными ядрами, зафиксированы в глобальной памяти перед доступом из других ядер.
Другие решения
Других решений пока нет …