Announcement

Collapse
No announcement yet.

Clang Gets Basic AMDGPU Toolchain Support

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

  • Clang Gets Basic AMDGPU Toolchain Support

    Phoronix: Clang Gets Basic AMDGPU Toolchain Support

    AMD's Tom Stellard has added a basic AMDGPU toolchain driver to the Clang compiler...

    http://www.phoronix.com/scan.php?pag...DGPU-Toolchain

  • #2
    if I compile a program with -mcpu=kaveri, then how to run it? for sure it's not ./a.out

    Comment


    • #3
      Originally posted by souenzzo View Post
      if I compile a program with -mcpu=kaveri, then how to run it? for sure it's not ./a.out
      I guess you will need to load it onto GPU with OpenCL or something similar.

      Comment


      • #4
        Originally posted by souenzzo View Post
        if I compile a program with -mcpu=kaveri, then how to run it? for sure it's not ./a.out
        You can use the HSA runtime to execute the program. I'm still working on getting everything upstream and open sourced, but the goal is to allow assembly programs or higher-level language programs to be compiled by clang to a binary format that the HSA runtime can interpret.

        Comment


        • #5
          Originally posted by tstellar View Post

          You can use the HSA runtime to execute the program. I'm still working on getting everything upstream and open sourced, but the goal is to allow assembly programs or higher-level language programs to be compiled by clang to a binary format that the HSA runtime can interpret.
          GCC can do it already?

          Comment


          • #6
            Yes, I think both gcc (Suse) and llvm (Tom) were doing this a few months ago -- we used the llvm paths as part of the all-open runtime that accompanied the request for upstreaming into 3.19.

            Comment


            • #7
              Can someone clarify my confusion please?

              I thought that all these changes are about getting clang to target the GPU during compile. Are they are about getting clang to produce software that can use the GPU during execution? Doesn't a clang "target" refer to "where it is compiled"?

              Comment


              • #8
                Originally posted by jaxxed View Post
                I thought that all these changes are about getting clang to target the GPU during compile. Are they are about getting clang to produce software that can use the GPU during execution?
                Correct. The purpose of this toolchain is to produce code which runs entirely on the GPU, eg for compiling compute kernels.

                Typical usage of the compiled code would be passing it to a "dispatch" operation which basically says "run a copy of this program in parallel for every element in this 4096 x 4096 element matrix". The dispatch operation could be something like a clEnqueueNDRangeKernel function in OpenCL or a cross-vendor AQL command in an HSA command queue.

                Originally posted by jaxxed View Post
                Doesn't a clang "target" refer to "where it is compiled"?
                AFAIK a clang "target" determines which instruction set is used when generating code (specified as an Architecture/Vendor/OS "triple"), so more like "what it is compiled for" than "where it is compiled". I imagine that by default the compiler would generate code for whatever host it is running on, so that might be where the confusion is coming from.

                Definitely not attempting to do the actual compilation on the GPU. Not yet anyways...

                BTW other compilers exist which produce mixed CPU and GPU code, typically using source code tags to identify sections that can be executed in parallel on the GPU. An example is the Kalmar C++ AMP toolchain by MCW:

                https://bitbucket.org/multicoreware/...pport%20Status

                (I believe Kalmar was previously called Clamp, when the focus was mostly c++ AMP running over openCL)
                Last edited by bridgman; 18 July 2015, 05:49 PM.

                Comment


                • #9
                  Thanks Bridge.

                  Comment

                  Working...
                  X