2013-03-13 1 views
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]; 
    } 
} 

답변

1

작업 항목은 0,1,2,3, ... 제 0,1,2,3 버퍼 인덱스를 읽을 것이다 (이것은 일반적으로 메모리 액세스를위한 최선의 경우이다) 병렬로, 그리고 1000100110021003, ... 등 병렬 처리.

커널 코드의 각 명령어는 모든 작업 항목에 의해 "병렬로"실행됩니다.

+0

아, 맞습니다. 나는 한 번에 하나의 커널을 읽고 생각했다. 다시 드로잉 보드 커널을 다시 설계 :) –

+0

그 get_global_size 동시에 동시에 생성되는 작업 항목 수를 같아야합니다. –

+0

예, 전체 크기 *는 "동시에 실행"하도록 예약 된 작업 항목 수입니다. 입력 버퍼는 그보다 훨씬 클 수 있으며 while 루프의 목적입니다. 전체 버퍼를 "get_global_size (0)"요소 청크로 덮어 두십시오. 끝날 때마다 각 작업 항목에는 데이터의 "자신의"슬라이스의 최소값이 있습니다. –