Abysmal OpenCL ImageSampling performance vs OpenGL TextureSampling

Question

Niedawno zaimportowałem volumeraycaster z OpenGL do OpenCL, który zmniejszył wydajność Raycastera o około 90 procent. Śledziłem spadek wydajności do funkcji próbkowania OpenCL, które są znacznie wolniejsze niż odpowiednie funkcje próbkowania OpenGL. Po usunięciu funkcji robienia zdjęć i funkcji próbkowania tekstur obie wersje Raycastera miały mniej więcej taką samą prędkość. W celu łatwego rozmieszczenia funkcji na różnych urządzeniach i wykluczenia niektórych głupich błędów w pozostałym kodzie RTs, napisałem mały test porównawczy, który porównuje prędkość próbkowania OpenCL do prędkości próbkowania OpenGL i przetestował ją na różnych komputerach ale OpenCL nadal miał około 10% wydajności OpenGL.

wzorca porównawczego OpenCL HostCode (przynajmniej najważniejszą część):

void OGLWidget::OCLImageSampleTest() 
{ 
    try 
    { 
    int size=8; 
    float Values[4*size*size*size]; 
    cl::Kernel kernel=cl::Kernel(program,"ImageSampleTest",NULL); 
    cl::ImageFormat FormatA(CL_RGBA,CL_FLOAT); 
    cl::Image3D CLImage(CLcontext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR ,FormatA,size,size,size,0,0,Values,NULL); 


    cl::ImageFormat FormatB(CL_RGBA,CL_UNSIGNED_INT8); 
    cl::Image2D TempImage(CLcontext, CL_MEM_WRITE_ONLY,FormatB,1024,1024,0,NULL,NULL); 


    kernel.setArg(0, CLImage); 
    kernel.setArg(1, TempImage); 



    cl::Sampler Samp; 
    Samp() = clCreateSampler(CLcontext(), CL_TRUE, CL_ADDRESS_REPEAT, CL_FILTER_LINEAR, NULL); 
    kernel.setArg(2, Samp); 

    QTime BenchmarkTimer=QTime(); 
    BenchmarkTimer.start(); 

    cl::KernelFunctor func = kernel.bind(queue, cl::NDRange(1024,1024), cl::NDRange(32,32)); 
    func().wait(); 

    int Duration = BenchmarkTimer.elapsed(); 
    printf("OCLImageSampleTest: %d ms \n", Duration); 
    } 
    catch (cl::Error& err) 
     { 
     std::cerr << "An OpenCL error occured, " << err.what() 
        << "\nError num of " << err.err() << "\n"; 
     return; 
     } 

} 

OpenCL jądra:

void kernel ImageSampleTest(read_only image3d_t CoordTexture, write_only image2d_t FrameBuffer, sampler_t smp) 
{ 
int Screenx = get_global_id(0); 
int Screeny = get_global_id(1); 

int2 PositionOnScreen=(int2)(Screenx,Screeny) ; 

float4 Testvec=(float4)(1,1,1,1); 
for(int i=0; i< 2000; i++) 
{ 
Testvec+= read_imagef(CoordTexture,smp, (float4)(0+0.00000001*i,0,0,0)); // i makes sure that the compiler doesn't unroll the loop 
} 

uint4 ToInt=(uint4)((uint) (Testvec.x), (uint) (Testvec.y) ,(uint)(Testvec.z),1); 
write_imageui ( FrameBuffer, PositionOnScreen, ToInt); 

} 

OpenGL FragmentShader na pełnym ekranie quad, który ma taką samą ilość fragmentów jako jądro OpenCL ma elementy pracy:

#version 150 
uniform sampler3D Tex; 
out vec4 FragColor; 

void main() 
{ 
FragColor=vec4(0,0,0,0); 
for(int i=0; i<2000; i++) 
{ 
FragColor+= texture(Tex,vec3(0+0.00000001*i,0,0),0); 
} 
} 

Ponadto już wypróbowałem e czynności, aby zwiększyć wydajność:

-Zmiana rozmiaru robocza: brak wzrostu wydajności

-Different Hardware: 280 GTX, GTX 580, niektóre karty Fermi Tessla, wszyscy mieli ten sam fatalne wyniki w OpenCL vs OpenGL

-Different formatów tekstur (bajty zamiast pływaków), różne wzory i różne rozmiary dostępu tekstury: brak wzrostu

-Korzystanie bufor zamiast obrazu dla danych i self napisane funkcja interpolacji trójliniowe pobierania próbek w jądrze języka CL: Zwiększono wydajność OpenCL o około 100%

-Używanie obrazu 2D // tekstura zamiast obrazu 3D // tekstura: zwiększyło to wydajność OpenCL o 100%, chociaż wydajność OpenGL w ogóle się nie zmieniła.

-Korzystanie „najbliższy” zamiast „interpolacji liniowej”: bez zmian wydajność

ta zostawiła mi zastanawiasz: Zrobiłem bardzo głupi błąd, który zmniejsza wydajność OpenCL? Dlaczego wydajność próbkowania OpenCL jest tak niska, chociaż powinna używać tego samego sprzętu teksturującego co OpenGL? Dlaczego moja złożona implementacja funkcji interpolacji trójliniowej jest szybsza niż implementacja sprzętowa? Jak zwiększyć wydajność próbkowania w OpenCL, aby móc korzystać z tej samej szybkości, co w OpenGL?

Answer 1

Podejrzewam, że jest jakiś problem z OpenCL w najnowszych sterownikach NVidii na niektórych kartach graficznych. Here i here to niektóre raporty na ich temat. Spróbuj powtórzyć test na GPU z innej rodziny.