Announcement

Collapse
No announcement yet.

AMD Radeon R600 GPU LLVM 3.3 Back-End Testing

Collapse
X
  • Filter
  • Time
  • Show
Clear All
new posts

  • AMD Radeon R600 GPU LLVM 3.3 Back-End Testing

    Phoronix: AMD Radeon R600 GPU LLVM 3.3 Back-End Testing

    One of the exciting features of LLVM 3.3 that is due out next month is the final integration of the AMD R600 GPU LLVM back-end. This LLVM back-end is needed for supporting Gallium3D OpenCL on AMD Radeon graphics hardware, "RadeonSI" HD 7000/8000 series support, and can optionally be used as the Radeon Gallium3D driver's shader compiler. In this article are some benchmarks of the AMD R600 GPU LLVM back-end from LLVM 3.3-rc1 when using several different AMD Radeon HD graphics cards and seeing how the LLVM compiler back-end affects the OpenGL graphics performance.

    http://www.phoronix.com/vr.php?view=18709

  • #2
    Gains are more noticeable in Unigine Heaven and Lightmarks. (Xonotic/Doom/Warsow aren't really shader limited).

    Comment


    • #3
      Should have been a 3-way comparison

      No offence, but this should have been a 3-way comparison : default shader compiler, llvm an Vadim Girlin's sb. (Sorry if I got the name wrong). Knowing that these different backends are runtime selectable, there is no excuse.

      Also, does anyone know if there are piglit results differences between all these backends? (I don't have the time to run it right now). In any case, it's nice to see this llvm based compiler work. More code sharing, that's always good.
      Serafean

      Comment


      • #4
        OpenCL

        Does this means fully working OpenCL is near ? I would really like to use Blender Cycles with OpenCL, even though I'm tempted to buy a Nvidia graphics card only for CUDA.

        Comment


        • #5
          Originally posted by wargames View Post
          Does this means fully working OpenCL is near ? I would really like to use Blender Cycles with OpenCL, even though I'm tempted to buy a Nvidia graphics card only for CUDA.
          We've got bfgminer working (minus a lock-up issue on some evergreens, possibly due to some flushing issues), and I believe that many of the Gimp GEGL operations are supported. Cycles may be in the cards, but last time I tried to compile their shaders (several months ago), they were not working, and looked like they broke some of the CL standard... but it was hard to tell, given that it was all CUDA code that was ported/translated through pre-processor macros.

          Comment


          • #6
            Originally posted by wargames View Post
            Does this means fully working OpenCL is near ? I would really like to use Blender Cycles with OpenCL, even though I'm tempted to buy a Nvidia graphics card only for CUDA.
            If you've got more than 1 PCIe slot you could just get the nvidia. I've done that before, using my HD5750 for gaming with a 8400GS for Physx.

            Comment


            • #7
              Originally posted by Veerappan View Post
              We've got bfgminer working (minus a lock-up issue on some evergreens, possibly due to some flushing issues), and I believe that many of the Gimp GEGL operations are supported. Cycles may be in the cards, but last time I tried to compile their shaders (several months ago), they were not working, and looked like they broke some of the CL standard... but it was hard to tell, given that it was all CUDA code that was ported/translated through pre-processor macros.
              Blender has no qualified OpenCL staff to make their code work correctly. They keep blaming AMD when it's on them.

              Comment


              • #8
                Originally posted by Serafean View Post
                No offence, but this should have been a 3-way comparison : default shader compiler, llvm and Vadim Girlin's sb. (Sorry if I got the name wrong). Knowing that these different backends are runtime selectable, there is no excuse.

                Also, does anyone know if there are piglit results differences between all these backends? (I don't have the time to run it right now). In any case, it's nice to see this llvm based compiler work. More code sharing, that's always good.
                Serafean
                Nitpick: That's a four-way comparison, since sb is not a shader compiler but a post-compile shader optimizer that can be used with both backends.

                Comment


                • #9
                  Originally posted by Marc Driftmeyer View Post
                  Blender has no qualified OpenCL staff to make their code work correctly. They keep blaming AMD when it's on them.
                  Yeah, right... but wait, the issue "does not appear with NVidia GPU OpenCL implementation neither on Intel/AMD CPU OpenCL implementations.": http://www.youtube.com/watch?v=LbEZ6OnpWHA . Btw AMD guys seem to be working on a fix: http://devgurus.amd.com/message/1285984

                  Comment


                  • #10
                    I've emerged the latest related packages from Gentoo's x11 overlay in order to enable opencl support on my 6970M using the radeon driver, and it all emerged cleanly. Is there a simple way to test whether the opencl support exists, or simply get info on what's enabled?

                    Thanks..

                    Comment


                    • #11
                      Originally posted by jasn View Post
                      I've emerged the latest related packages from Gentoo's x11 overlay in order to enable opencl support on my 6970M using the radeon driver, and it all emerged cleanly. Is there a simple way to test whether the opencl support exists, or simply get info on what's enabled?

                      Thanks..
                      https://wiki.archlinux.org/index.php/GPGPU

                      perhaps cuda_memtest gives some meaningful output ?

                      Comment


                      • #12
                        Originally posted by kernelOfTruth View Post
                        perhaps cuda_memtest gives some meaningful output ?
                        Thanks for the suggestion. However, when I added sekyfsr's overlay in order to emerge cuda_memtest, I realized that on my system this would require emerging gcc 4.4, and 4.6, and the cuda toolkit, before cuda_memtest. I'm hoping that there's a simpler method of simply determining if it's working. I next tried emerging the phoronix-test-suite-4.4.1, and running both the pyopencl test, and the opencl test suite, and both came back with failure messages indicating that I don't have opencl working on my system, ("The test run did not produce a result.", and "This test failed to run properly.")

                        So after checking out Tom Stellard's blog post here, I'm now wondering if the code changes that he mentions there have been pushed to the normal master branches of the llvm, mesa, and libclc, respective repositories. The Gentoo x11 overlay libclc live ebuild does pull from his git repository, but doesn't specify the bfgminer branch, and the llvm and mesa live ebuilds don't pull from his git repositories at all, but rather the normal master branch repositories, (as one would expect).

                        I'm going to try and edit those live ebuilds after adding them to my local overlay, and see if I can pull from Tom's bfgminer branches of those repositories and re-emerge llvm, mesa, and libclc. I'm hoping that the changes he mentions that are required for clang, ("checkout r179204 from the clang tree"), have been rolled into the current version in their repository. I think it's 182332. If I can get that far, I'll retry running the phoronix tests and see if there's any difference.

                        Anyone else has some advice please feel free to pass it along.

                        Thanks..

                        Comment


                        • #13
                          Originally posted by jasn View Post
                          Thanks for the suggestion. However, when I added sekyfsr's overlay in order to emerge cuda_memtest, I realized that on my system this would require emerging gcc 4.4, and 4.6, and the cuda toolkit, before cuda_memtest. I'm hoping that there's a simpler method of simply determining if it's working. I next tried emerging the phoronix-test-suite-4.4.1, and running both the pyopencl test, and the opencl test suite, and both came back with failure messages indicating that I don't have opencl working on my system, ("The test run did not produce a result.", and "This test failed to run properly.")
                          If you need to verify that you have OpenCL installed correctly, you can use these simple demos: http://cgit.freedesktop.org/~tstellar/opencl-example/ I usually use the hello_world for this.

                          So after checking out Tom Stellard's blog post here, I'm now wondering if the code changes that he mentions there have been pushed to the normal master branches of the llvm, mesa, and libclc, respective repositories. The Gentoo x11 overlay libclc live ebuild does pull from his git repository, but doesn't specify the bfgminer branch, and the llvm and mesa live ebuilds don't pull from his git repositories at all, but rather the normal master branch repositories, (as one would expect).

                          I'm going to try and edit those live ebuilds after adding them to my local overlay, and see if I can pull from Tom's bfgminer branches of those repositories and re-emerge llvm, mesa, and libclc. I'm hoping that the changes he mentions that are required for clang, ("checkout r179204 from the clang tree"), have been rolled into the current version in their repository. I think it's 182332. If I can get that far, I'll retry running the phoronix tests and see if there's any difference.

                          Anyone else has some advice please feel free to pass it along.

                          Thanks..
                          All of the code in my bfgminer branches has been pushed upstream, so everything should work fine with the gentoo overlay.

                          Comment


                          • #14
                            Thanks Tom..

                            Originally posted by tstellar View Post
                            If you need to verify that you have OpenCL installed correctly, you can use these simple demos: http://cgit.freedesktop.org/~tstellar/opencl-example/ I usually use the hello_world for this.
                            (Proving that it's dangerous to give a compiler to someone who's never written a program in his life), I tried compiling the programs in your opencl-example repository, but I'm getting no joy. Here's the output from issuing a make command from the directory;
                            Code:
                            make
                            cc -g   -c -o hello_world.o hello_world.c
                            cc -g   -c -o cl_simple.o cl_simple.c
                            cl_simple.c: In function 'clSimpleCreateBuffer':
                            cl_simple.c:80:24: warning: format '%u' expects argument of type 'unsigned int', but argument 5 has type 'cl_mem_flags' [-Wformat]
                            cl_simple.c:80:24: warning: format '%u' expects argument of type 'unsigned int', but argument 6 has type 'size_t' [-Wformat]
                            cc -g   -c -o cl_util.o cl_util.c
                            gcc -o hello_world hello_world.o cl_simple.o cl_util.o -L/usr/local/lib -lOpenCL
                            /usr/lib/gcc/x86_64-pc-linux-gnu/4.7.3/../../../../lib64/libOpenCL.so: undefined reference to `llvm::DIBuilder::createImportedModule(llvm::DIScope, llvm::DINameSpace, unsigned int)'
                            collect2: error: ld returned 1 exit status
                            make: *** [hello_world] Error 1
                            Please feel free to point out the obvious to me..

                            Originally posted by tstellar View Post
                            All of the code in my bfgminer branches has been pushed upstream, so everything should work fine with the gentoo overlay.
                            Again, thanks for not only the confirmation, but thanks for all the work as well.

                            Jason

                            Comment


                            • #15
                              Thank you Tom..

                              (for the work and reply)

                              I was able to compile the opencl-example programs after updating my system today which included updating the latest git versions of clang, llvm, and mesa. The output does seem to indicate that opencl is indeed working on my system;
                              Code:
                              ./hello_world
                              There are 1 platforms.
                              There are 1 GPU devices.
                              clCreateContext() succeeded.
                              clCreateCommandQueue() succeeded.
                              clCreateProgramWithSource() suceeded.
                              clBuildProgram() suceeded.
                              clCreateKernel() suceeded.
                              clCreateBuffer() succeeded.
                              clSetKernelArg() succeeded.
                              clEnqueueNDRangeKernel() suceeded.
                              clEnqueueReadBuffer() suceeded.
                              pi = 0.000000
                              Thanks again..

                              Jason

                              Comment

                              Working...
                              X