異常なOpenCL ImageSamplingの性能とOpenGLのテクスチャサンプリング

Question

私は最近、OpenGLからOpenCLへvolumeraycasterを移植しました。これはレイキャスターのパフォーマンスを約90％低下させました。私は、それぞれのOpenGLテクスチャサンプリング関数よりもはるかに遅いOpenCLの画像サンプリング関数の性能低下を追跡しました。画像サンプリング関数とテクスチャサンプリング関数を削除することにより、両方のレイキャスターの実装にはほぼ同じ速度がありました。 異なるハードウェアの機能を簡単にベンチマークし、残りのRTコードの間違いを排除するために、私はOpenCLサンプリング速度とOpenGLサンプリング速度を比較し、異なるマシンでテストした小さなベンチマークを作成しましたOpenCLはまだOpenGLの性能の約10％しか持っていませんでした。

ベンチマークのOpenCLのHostCode（それの少なくとも最も重要な部分）：

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カーネル：

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 OpenCLカーネルは作業項目を持っています：

#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); 
} 
} 

さらに私はすでに電子パフォーマンスを向上するには、次

ワークグループサイズ-changing：なし性能向上を

-Differentハードウェア：280 GTX 580 GTX、いくつかのフェルミTesslaカードは、それらのすべては、OpenGL対のOpenCLで同じ底抜けの性能を持っていました

-Differentテクスチャフォーマット（代わりに、フロートのバイト）、異なるアクセスパターンと異なるテクスチャサイズ：サンプリングデータ用画像の代わりにバッファを-Usingない増加

及び自己書き込まトリリニア補間関数CLカーネルにおけるOpenCLのパフォーマンスの向上約100％

-3D画像ではなくテクスチャ//テクスチャ：OpenGLの性能が全く変わっていないのに、OpenCLの性能が100％向上しました。

「最も近い」の代わりに「線形」補間の-Using：これは不思議私に残さ

ない性能変化： 私はOpenCLの性能を低下させる非常に愚かな間違いをしましたか？ OpenCLと同じテクスチャハードウェアを使用する必要がありますが、OpenCLサンプリングのパフォーマンスがそれほど低いのはなぜですか？ 私の複雑な三線式補間関数の実装は、ハードウェアの実装よりも速いのはなぜですか？ OpenGLのサンプリング性能を向上させて、OpenGLの場合と同じ速度にすることができますか？

Answer 1

一部のビデオカードで最新のNVidiaドライバでOpenCLに問題があると思われます。 Hereおよびhereは、これらの報告書です。別のファミリのGPUでテストを繰り返してみてください。