2012-09-21 1 views
1

각 작업 그룹이 로컬 메모리에 결과 벡터를 생성하는 OpenCL 커널이 있습니다. 그런 다음 나중에 호스트로 검색하기 위해 이러한 결과를 모두 전역 메모리로 합산해야합니다. 본질적으로OpenCL에서 로컬 메모리에서 전역 메모리로 결과를 집계하는 방법

//1st thread in each workgroup initializes local buffer 
if(get_local_id(0) == 0){ 
    for(i=0; i<HYD_DIM; i++){ 
     pressure_Local[i] = (float2){1.0f, 0.0f}; 
    } 
} 

//wait for all workgroups to finish accessing any memory 
barrier(CLK_GLOBAL_MEM_FENCE | CLK_LOCAL_MEM_FENCE); 

/// sum all the results into global storage 
for(i=0; i<get_num_groups(0); i++){ 

    //1st thread in each workgroup writes the group's local buffer to global memory 
    if(i == get_group_id(0) && get_local_id(0) == 0){ 
     for(j=0; j<HYD_DIM; j++){ 
      pressure_Global[j] += pressure_Local[j]; 
      // barrier(CLK_GLOBAL_MEM_FENCE); 
     } 
    } 

    //flush global memory buffers: 
    barrier(CLK_GLOBAL_MEM_FENCE); 
} 

, 나는 (내 경우에는 128) 작업 그룹의 수와 동일하게 전역 메모리에서 벡터의 모든 요소를 ​​기다리고 있었다 :
이를 테스트하기 위해, 나는 다음과 같은 커널 코드를 만들었습니다. 실제로 그들은 일반적으로 60에서 70 사이의 차이를 보이며, 결과는 실행마다 바뀝니다.
누락 된 부분이나 올바르게 처리하는 방법을 누군가에게 알려 줄 수 있습니까?

답변

2

다른 작업 그룹간에 opencl을 동기화 할 수 없습니다. CLK_GLOBAL_MEM_FENCE는 그렇게 작동하지 않습니다. 작업 그룹에 의해 액세스되는 메모리 조작의 순서가 유지된다는 것만을 보장합니다. OCL 1.2 spec의 "6.12.8 동기화 기능"절을 참조하십시오.

각 작업 그룹마다 다른 전역 메모리 블록을 사용하여 문제를 해결할 수 있습니다. 데이터를 전역에 기록하면 커널이 완료됩니다. 그런 다음 데이터를 단일 블록으로 줄이려면 다른 커널이 전역의 데이터를 읽고 다른 결과 블록과 병합하도록 할 수 있습니다. 원하는만큼 많은 병합 계층을 수행 할 수 있지만 최종 병합은 단일 작업 그룹에 의해 수행되어야합니다.

gpu/opencl 감소 알고리즘을 검색하십시오. 괜찮은 것부터 시작하겠습니다. Case Study: Simple Reductions

관련 문제