0
OpenCL을 절감을 이해하는 도움이 필요합니다.나는 다음과 같은 메모를 읽고있다
while (global_index < length) .... global_index += get_global_size(0)
필자는 순차적으로 레이아웃 된 전역 저장소의 데이터를 읽는 것이 더 현명하다고 역설했습니다. k, k + 1, k + 2에서의 데이터 읽기 의미는 k + 1000, k + 2000, k + 3000을 읽는 것이 더 빠릅니다. 이것은 global_index + = get_global_size (0)라고 할 때 그들이하는 일이 아닌가?
__kernel
void reduce(__global float* buffer,
__local float* scratch,
__const int length,
__global float* result) {
int global_index = get_global_id(0);
float accumulator = INFINITY;
// Loop sequentially over chunks of input vector
while (global_index < length) {
float element = buffer[global_index];
accumulator = (accumulator < element) ? accumulator : element;
global_index += get_global_size(0);
}
// Perform parallel reduction
int local_index = get_local_id(0);
scratch[local_index] = accumulator;
barrier(CLK_LOCAL_MEM_FENCE);
for(int offset = get_local_size(0)/2;
offset > 0;
offset = offset/2) {
if (local_index < offset) {
float other = scratch[local_index + offset];
float mine = scratch[local_index];
scratch[local_index] = (mine < other) ? mine : other;
}
barrier(CLK_LOCAL_MEM_FENCE);
}
if (local_index == 0) {
result[get_group_id(0)] = scratch[0];
}
}
아, 맞습니다. 나는 한 번에 하나의 커널을 읽고 생각했다. 다시 드로잉 보드 커널을 다시 설계 :) –
그 get_global_size 동시에 동시에 생성되는 작업 항목 수를 같아야합니다. –
예, 전체 크기 *는 "동시에 실행"하도록 예약 된 작업 항목 수입니다. 입력 버퍼는 그보다 훨씬 클 수 있으며 while 루프의 목적입니다. 전체 버퍼를 "get_global_size (0)"요소 청크로 덮어 두십시오. 끝날 때마다 각 작업 항목에는 데이터의 "자신의"슬라이스의 최소값이 있습니다. –