Звоните многократно get_global_id () по сравнению с сохранением результат в локальной переменной?
Это - вероятно, глупый вопрос, но: Насколько дорогой это для вызова некоторых get_* функция в OpenCL-ядрах? Лучше сохранить результат для будущего использования в некоторой локальной переменной или вызвать желаемую функцию каждый раз, когда этому было нужно?
Или это зависит платформа?
PS я думаю, cuda, решает его лучше с различными threadIdx переменными.
1 ответ
Я думаю, что это должно быть свободно для всех архитектур GPU. Он должен быть заменен соответствующим аппаратным регистром или константой в банке кэша.
Компилятор также может выполнять распространение константы на него. Вы можете проверить сами, используя AMD Stream Analyser:
OpenCL:
__kernel
void testKernel(__global uint * uintArray)
{
uint threadId = get_global_id(0);
uintArray[threadId] = 0xbaadf00d;
}
Radeon HD 5870 (Cypress) assembly:
0 ALU: ADDR(32) CNT(10) KCACHE0(CB0:0-15) KCACHE1(CB1:0-15)
0 x: MOV R1.x, (0xBAADF00D, -0.001327039325f).x
t: MULLO_INT ____, R1.x, KC0[1].x
1 x: ADD_INT ____, R0.x, PS0
2 w: ADD_INT ____, PV1.x, KC0[6].x
3 z: LSHL ____, PV2.w, (0x00000002, 2.802596929e-45f).x
4 y: ADD_INT ____, KC1[0].x, PV3.z
5 x: LSHR R0.x, PV4.y, (0x00000002, 2.802596929e-45f).x
01 MEM_RAT_CACHELESS_STORE_RAW: RAT(1)[R0].x___, R1, VPM
Здесь get_global_id(0) отображается на константное значение банка кэша KC0[1].x.
Итак, для ответа на ваш вопрос я бы использовал наиболее читаемую форму.