OpenCl의 스칼라 및 벡터 커널

Question

OpenCl을 사용하는 미니 분자 동역학 응용 프로그램에 대한이 문서 http://eprints.dcs.warwick.ac.uk/1694/1/miniMD_opencl.pdf을 읽는 중입니다. 코드는 여기에 있습니다.
커널을 구현하는 방법에 익숙해졌습니다. 내가 이해하지 못하는 것은이

#if defined(SCALAR_KERNELS) 
__kernel void f_clear(
    __global float* f, 
    __const int nall) { 

    for (unsigned i = get_global_id(0)+1; i <= nall; i += get_global_size(0)) { 
     const int i4 = i << 2; 
     f[i4+0] = 0.0f; 
     f[i4+1] = 0.0f; 
     f[i4+2] = 0.0f; 
     f[i4+3] = 0.0f; 
    } 

} 
#elif defined(VECTOR_KERNELS) 
__kernel __attribute__((vec_type_hint(float4))) 
void f_clear(
    __global float4* f, 
    __const int nall) { 

    const float4 zeroes = (float4) (0.0f, 0.0f, 0.0f, 0.0f); 
    for (unsigned i = get_global_id(0)+1; i <= nall; i += get_global_size(0)) { 
     f[i] = zeroes; 
    } 

} 
#endif 

가 VECTOR_KERNELS을 가정하고, SCALAR_KERNELS는 GPU와 MIC 장치 있지만 확실하지에 해당합니다.
MIMD SIMD 명령어 또는 멀티 코어 및 벡터 프로그래밍과 관련이 있습니까?
또한 현재 벡터 형식을 사용하는 것이 큰 이점이 있습니까?
마지막으로 루프에 대한 두 가지와 그 목적을 실제로 파악할 수 없습니다.
왜 f[get_global_id(0)]을 보내지 않으십니까?
감사합니다.
Éric.

Answer 1

AMD, ATI 및 Intel과 같은 일부 기기는 벡터 유형을 지원하는 데 정말로 유용합니다. 이러한 벡터는 SIMD이며 가능한 경우 사용하는 것이 더 빠릅니다. NVIDIA는 OpenCL에서 벡터를 지원할 때 매우 좋지 않습니다 (적어도 테스트 한 모든 것).

두 루프 모두 크기가 nall 인 글로벌 메모리 덩어리를 지우는 것으로 나타납니다.

Answer 2

스칼라와 벡터는 OpenCL에서 동일한 작업을 수행하는 다른 방법 일뿐입니다. 그러나 벡터는 컴파일러 (CPU 또는 GPU/FPGA)에서 더 잘 최적화되어야하므로으로 갈 수 있습니다. 이 방법으로 컴파일러는 자연스럽게 SIMD 유닛을 완성 할 수 있습니다. 그래서, 가능하고 쉬운 당신을 위해, 그들을 사용하십시오.

Austin이 말한 것처럼 두 루프 모두 전역 메모리 크기 인 nall을 정리합니다.

그러나 코드를 보면 매우 불충분합니다. 동일한 작업 그룹의 작업 항목은 완전히 다른 전역 메모리 영역에 액세스하여 병합을 해제합니다. 그것은 (당신이 말한대로) 단지에 의해 훨씬 더 좋을 것이다 :

__kernel __attribute__((vec_type_hint(float4))) 
void f_clear(
    __global float4* f) { 
    f[get_global_id(0)] = (float4) (0.0f, 0.0f, 0.0f, 0.0f); 
} 

그리고 적절한 글로벌 크기 (global_size = nall)이 커널을 시작하고 컴파일러가 로컬 작업 그룹 크기를 결정할 수 있습니다.

추신 : 나는 그것을해야 할 것입니다, 나는 clEnqueueWriteBuffer를 호출하고 CPU에서 메모리를 청소하는 것을 선호합니다. 다른 커널 실행과 병행하여 수행 할 수 있기 때문입니다.