My<cod/>.net
Ограничения на размер grid в Nvidia CUDA при двухмерной сетке?
Просмотров: 336
Ответов: 1
Всем привет. Возможно, мой вопрос покажется многим на первый взгляд очевидным, но я бы всё-таки призвал вас не считать его идиотским до тех пор, пока не прочтёте до конца.
Итак, в чём же суть вопроса. Как известно из документации CUDA, размер сетки запускаемого ядра имеет ограничения, которые зависят от конкретного устройства. В большинстве современных видеокарт ограничение равно 65535x65535x1. На моих видеокартах g210m и 8800gt это именно так, я проверял. Но в этом месте я встретил довольно странную вещь — в моей программе по какой-то неизвестной мне причине невозможно запустить ядро, которое бы имело размерность (по нитям) больше 5808x5808 (это число может быть меньше в зависимости от размера блока, я написал строгий максимум) или же больше 264х264 (если измерять в блоках) — и последнее число неизменно. Как только количество запускаемых блоков переваливает за 265х265, ядро запускается, отрабатывает, но в качестве результата всегда выдаёт ноль.
Дебаггер от Nvidia Nsight молчит, никаких ошибок не выбрасывается, профайлер выдаёт результаты работы, в которых ядро запускается. Ограничение всплывает на всех видеокартах, на которых я запускал программу — в сумме на 8 разных моделях (8400m g, 8800gt, 9600gso, 8500gt, 9600gt, ION, g210m, gf9300)
Так вот всё это наводит меня на мысль, что есть ограничения не только на размерность сетки, но и на суммарное количество нитей в сетке (ведь на количество нитей в блоке ограничение есть — почему бы и тут ему не быть). Только вот ни официальная документация, ни учебник Борескова/Харлмова, ни best practices guide ничего не этот счёт не говорят — просто говорят, что есть ограничения, уже озвученные в самом начале вопроса.
Поскольку копаюсь я с этим примерно по часа два в день на протяжении уже недели, и никакого прогресса нет, я прошу помощи — куда копать? Любые замечания приветствуются, если нужно сделать какие-то уточнения — скажите
Только что проверил. Мне не удалось повторить вашу проблему.
У меня GTX470.
Итак. Написал ядро:
__global__ void testKernel( int* g_odata)
{
if(threadIdx.x==0)
{
g_odata[2*(blockIdx.y*gridDim.x+blockIdx.x)] = blockIdx.y;
g_odata[2*(blockIdx.y*gridDim.x+blockIdx.x)+1] = blockIdx.x;
}
}
Запустил его на 8192х8192 блоков и 1024 треда(в ваших видяхах максимум 512 тредов в блоке, на ферми 1024):
dim3 grid( 8192, 8192, 1);
dim3 threads( 1024, 1, 1);
testKernel<<< grid, threads, 0 >>>( d_odata);
Естественно выделил память и т.п.
И получил последним элементом массива: 8191x8191.
На больших числах не тестировал, потому что память кончается :( Надо уже какую-то логику реализовывать.
А вообще не понятно откуда у вас эти не круглые значения 265, 264?
__global__ void testKernel( int* g_odata)
{
if(threadIdx.x==0)
{
g_odata[2*(blockIdx.y*gridDim.x+blockIdx.x)] = blockIdx.y;
g_odata[2*(blockIdx.y*gridDim.x+blockIdx.x)+1] = blockIdx.x;
}
}
dim3 grid( 8192, 8192, 1);
dim3 threads( 1024, 1, 1);
testKernel<<< grid, threads, 0 >>>( d_odata);