54 votes

Performances abyssales de l'échantillonnage d'images OpenCL par rapport à l'échantillonnage de textures OpenGL

J'ai récemment porté mon volumeraycaster d'OpenGL à OpenCL, ce qui a diminué les performances du raycaster d'environ 90%. J'ai attribué cette baisse de performance aux fonctions d'échantillonnage d'images d'OpenCL, qui sont beaucoup plus lentes que les fonctions d'échantillonnage de textures respectives d'OpenGL. En supprimant les fonctions d'échantillonnage d'images et les fonctions d'échantillonnage de texture, les deux implémentations de raycaster avaient à peu près la même vitesse. Afin d'évaluer facilement les fonctions sur différents matériels, et pour exclure quelques erreurs stupides dans le reste de mon code RTs, j'ai écrit un petit benchmark qui compare la vitesse d'échantillonnage d'OpenCL à la vitesse d'échantillonnage d'OpenGL et je l'ai testé sur différentes machines mais OpenCL n'avait toujours qu'environ 10 % de la performance d'OpenGL.

Le HostCode OpenCL du benchmark (du moins la partie la plus importante) :

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

}

Noyau 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 ); 

}

FragmentShader OpenGL pour un quad plein écran qui a le même nombre de fragments que le noyau OpenCL a d'éléments de travail :

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

En outre, j'ai déjà essayé ce qui suit pour augmenter les performances :

-modification de la taille du groupe de travail : pas d'augmentation des performances

-Différents matériels : 280 GTX, 580 GTX, certaines cartes Fermi Tessla, tous avaient les mêmes performances abyssales en OpenCL et OpenGL.

-Différents formats de texture (octets au lieu de flottants), différents schémas d'accès et différentes tailles de texture : aucune augmentation.

-Utilisation d'un tampon au lieu d'une image pour les données et d'une fonction d'interpolation trilinéaire auto écrite pour l'échantillonnage dans le noyau CL : Augmentation des performances d'OpenCL d'environ 100 %.

-Utilisation d'une image/texture 2D au lieu d'une image/texture 3D : Cela a augmenté les performances d'OpenCL de 100 % alors que les performances d'OpenGL n'ont pas changé du tout.

-Utilisation de l'interpolation "la plus proche" au lieu de l'interpolation "linéaire" : Pas de changement de performance

Je me suis alors demandé : Ai-je fait une erreur très stupide qui diminue les performances d'OpenCL ? Pourquoi la performance d'échantillonnage d'OpenCL est-elle si faible, alors qu'elle devrait utiliser le même matériel de texture qu'OpenGL ? Pourquoi l'implémentation de ma fonction d'interpolation trilinéaire complexe est-elle plus rapide que son implémentation matérielle ? Comment puis-je augmenter les performances d'échantillonnage d'OpenCL afin d'obtenir la même vitesse qu'avec OpenGL ?

3voto

FractalizeR Points 12887

Je pense qu'il y a un problème avec OpenCL dans les derniers pilotes NVidia sur certaines cartes vidéo. Ici y aquí Il existe des rapports à ce sujet. Essayez de répéter le test sur un GPU d'une autre famille.

Prograide.com

Prograide est une communauté de développeurs qui cherche à élargir la connaissance de la programmation au-delà de l'anglais.
Pour cela nous avons les plus grands doutes résolus en français et vous pouvez aussi poser vos propres questions ou résoudre celles des autres.

Powered by:

X