search for: spir64

...lvm/ADT/Triple.h (revision 212686) +++ include/llvm/ADT/Triple.h (working copy) @@ -76,7 +76,8 @@ le32, // le32: generic little-endian 32-bit CPU (PNaCl / Emscripten) amdil, // amdil: amd IL spir, // SPIR: standard portable IR for OpenCL 32-bit version - spir64 // SPIR: standard portable IR for OpenCL 64-bit version + spir64, // SPIR: standard portable IR for OpenCL 64-bit version + kalimba // Kalimba: generic kalimba }; enum VendorType { UnknownVendor, @@ -89,7 +90,8 @@ Freescale, IBM, ImaginationTechnol...

...ude/llvm/ADT/Triple.h (revision 212519) +++ include/llvm/ADT/Triple.h (working copy) @@ -76,24 +76,9 @@ le32, // le32: generic little-endian 32-bit CPU (PNaCl / Emscripten) amdil, // amdil: amd IL spir, // SPIR: standard portable IR for OpenCL 32-bit version - spir64 // SPIR: standard portable IR for OpenCL 64-bit version + spir64, // SPIR: standard portable IR for OpenCL 64-bit version + kalimba // Kalimba: generic kalimba }; - enum SubArchType { - NoSubArch, - - ARMSubArch_v8, - ARMSubArch_v7, - ARMSubArch_v7em, - ARMSub...

On 7/9/14, 12:33 PM, Eric Christopher wrote: > Any reason why you deleted code that isn't related? > > -eric > >> - enum SubArchType { >> - NoSubArch, >> - >> - ARMSubArch_v8, >> - ARMSubArch_v7, >> - ARMSubArch_v7em, >> - ARMSubArch_v7m, >> - ARMSubArch_v7s, >> - ARMSubArch_v6, >> -

...ts (such as specifying address spaces etc.). I guess I just need to ensure that I am sticking to LLVM 3.2 and the SPIR specified subset+ metadata annotations etc. Anyway, I was confused partly because there are references to a SPIR backend in LLVM code (eg: include/llvm/ADT/Triple.h lists spir and spir64 as possible architectures) rahui On Fri, Sep 5, 2014 at 2:55 PM, Tim Northover <t.p.northover at gmail.com> wrote: > Hi Rahul, > > On 5 September 2014 19:21, Rahul Garg <rahulgarg44 at gmail.com> wrote: >> I am looking into generating SPIR from within a JIT compiler for...

...a “compatible” data layout is, or you considered that it is up to the OpenCL vendor to figure out what will work or not, with the drawback that any LLVM update can break its use case. + For OpenCL, LLVM IR translated from SPIR-V has specific data layouts, which are the data layouts for target spir/spir64. OpenCL vendor’s target data layout are assumed to be consistent with them. For OpenCL kernels, there is implicit data layout dependence when compiling the source to LLVM. Since SPIR-V is for common OpenCL platforms, a common data layout accepted by different OpenCL vendors is required. We choos...

...ut is, or you considered that it is up to the OpenCL vendor to > figure out what will work or not, with the drawback that any LLVM update > can break its use case. > > + For OpenCL, LLVM IR translated from SPIR-V has specific data layouts, > which are the data layouts for target spir/spir64. OpenCL vendor’s target > data layout are assumed to be consistent with them. > > > > For OpenCL kernels, there is implicit data layout dependence when > compiling the source to LLVM. Since SPIR-V is for common OpenCL platforms, > a common data layout accepted by different Open...

...rst step is to of course build LLVM with SPIR backend support and I am currently using LLVM 3.3. For reference, the host system is a x86-64 Linux. I first tried building LLVM 3.3 without specifying any targets, and it built many targets including aarch64, arm, x86, x86_64, nvptx etc but no spir or spir64 targets. Next, I tried to explicitly configure targets as follows: --enable-targets=x86_64,spir,spir64 but I got "unrecognized target spir". Any idea how to proceed? Do I need to apply some patches to LLVM? (Note that Clang is not involved, this is purely a LLVM question). rahul

...is up to the OpenCL vendor to > figure out what will work or not, with the drawback that any > LLVM update can break its use case. > > + For OpenCL, LLVM IR translated from SPIR-V has specific data > layouts, which are the data layouts for target spir/spir64. > OpenCL vendor’s target data layout are assumed to be > consistent with them. > > For OpenCL kernels, there is implicit data layout dependence > when compiling the source to LLVM. Since SPIR-V is for common > OpenCL platforms, a common dat...

Hi all, According to the SPIR 2.0 spec[1], the name of OpenCL builtins are mangled. However, when I compile OpenCl code with Clang 3.9 with the "spir64-unknown-unknown" target, Clang generates IR without mangling the builtins, e.g. for: __kernel void input_zip_int(__global int *in0) { *in0 = get_global_id(0); } clang generates: define spir_kernel void @input_zip_int(i32 addrspace(1)* nocapture %in0) local_unnamed_addr #0 !kernel_arg_addr...

...a “compatible” data layout is, or you considered that it is up to the OpenCL vendor to figure out what will work or not, with the drawback that any LLVM update can break its use case. + For OpenCL, LLVM IR translated from SPIR-V has specific data layouts, which are the data layouts for target spir/spir64. OpenCL vendor’s target data layout are assumed to be consistent with them. For OpenCL kernels, there is implicit data layout dependence when compiling the source to LLVM. Since SPIR-V is for common OpenCL platforms, a common data layout accepted by different OpenCL vendors is required. We choose...

...figure out what will work or not, with the drawback that any >>> LLVM update can break its use case. >>> >>> + For OpenCL, LLVM IR translated from SPIR-V has specific data >>> layouts, which are the data layouts for target spir/spir64. >>> OpenCL vendor’s target data layout are assumed to be >>> consistent with them. >>> >>> For OpenCL kernels, there is implicit data layout dependence >>> when compiling the source to LLVM. Since SPIR-V is for common...

...Sam) <Yaxun.Liu at amd.com> > *Subject:* builtins name mangling in SPIR 2.0 > > > > Hi all, > > > > According to the SPIR 2.0 spec[1], the name of OpenCL builtins are mangled. > > > > However, when I compile OpenCl code with Clang 3.9 with the > "spir64-unknown-unknown" target, Clang generates IR without mangling the > builtins, e.g. for: > > > > __kernel void input_zip_int(__global int *in0) { > > *in0 = get_global_id(0); > > } > > > > clang generates: > > > > define spir_kernel void @inpu...

Hi, I’ve recently encountered an issue where the `instcombine` pass replaces an `llvm.memcpy` between two distinct address spaces with an `addrspacecast` instruction. As an example, see the trivial OpenCL kernel attached. I’m compiling like this: clang -cc1 -triple spir64-unknown-unknown -x cl -O0 -emit-llvm array_init.cl -o before.ll This yields an `llvm.memcpy` to copy the array initialiser data from the global variable (in `addrspace(2)`) to the `alloca` result (in `addrspace(0)`). I then apply the `instcombine` pass via: opt -S -instcombine before.ll -o a...

...yei at intel.com> wrote: > Hi Ankur,**** > > SPIR generation is not yet fully supported in Clang. I’m working on this > task, but there are still some missing parts.**** > > The correct way to generate SPIR by Clang is adding the “-triple > spir-unknown-unknown” or “-triple spir64-unknown-unknown” command line > options when running clang.**** > > ** ** > > Thanks**** > > Guy Benyei**** > > [image: email_signature_guy_new2]**** > > ** ** > > *From:* cfe-dev-bounces at cs.uiuc.edu [mailto:cfe-dev-bounces at cs.uiuc.edu] *On > Behal...

...sts.llvm.org>>; Liu, Yaxun (Sam) <Yaxun.Liu at amd.com<mailto:Yaxun.Liu at amd.com>> Subject: builtins name mangling in SPIR 2.0 Hi all, According to the SPIR 2.0 spec[1], the name of OpenCL builtins are mangled. However, when I compile OpenCl code with Clang 3.9 with the "spir64-unknown-unknown" target, Clang generates IR without mangling the builtins, e.g. for: __kernel void input_zip_int(__global int *in0) { *in0 = get_global_id(0); } clang generates: define spir_kernel void @input_zip_int(i32 addrspace(1)* nocapture %in0) local_unnamed_addr #0 !kernel_arg_addr...

> On May 19, 2015, at 7:32 PM, Sean Silva <chisophugis at gmail.com> wrote: > > > > On Tue, May 19, 2015 at 4:05 PM, Owen Anderson <resistor at mac.com <mailto:resistor at mac.com>> wrote: > >> On May 19, 2015, at 9:48 AM, Neil Henning <llvm at duskborn.com <mailto:llvm at duskborn.com>> wrote: >> >> The 'backend' in

...ompilerInvocation.cpp index 619ea9c..d02d873 100644 --- a/clang/lib/Frontend/CompilerInvocation.cpp +++ b/clang/lib/Frontend/CompilerInvocation.cpp @@ -2437,6 +2437,13 @@ bool CompilerInvocation::CreateFromArgs(CompilerInvocation &Res, if (Arch == llvm::Triple::spir || Arch == llvm::Triple::spir64) { Res.getDiagnosticOpts().Warnings.push_back("spir-compat"); } + + // If there will ever be e.g. "LangOpts.C", replace "LangOpts.C11 || LangOpts.C99" with "LangOpts.C" on the next line. + if ( (LangOpts.C11 || LangOpts.C99 || LangOpts.CPlusPl...

...ve recently encountered an issue where the `instcombine` pass replaces an `llvm.memcpy` between two distinct address spaces with an `addrspacecast` instruction. >> >> As an example, see the trivial OpenCL kernel attached. I’m compiling like this: >> >> clang -cc1 -triple spir64-unknown-unknown -x cl -O0 -emit-llvm array_init.cl -o before.ll >> >> This yields an `llvm.memcpy` to copy the array initialiser data from the global variable (in `addrspace(2)`) to the `alloca` result (in `addrspace(0)`). >> >> I then apply the `instcombine` pass via: &gt...

...decoration for alignment, so a user could decorate a type to specify a required alignment (which I would have thought in turn would become part of the data layout). Also if we are using a non-logical addressing mode the data layout would have a different pointer width specified (similar to the SPIR/SPIR64 targets in Clang at present). I'll bring it up with the SPIR group at Khronos what the expected behaviour of the alignment decoration is in this context, but at present I would say it would be legal for an LLVM module that is being turned into SPIR-V to have a user-defined data layout. Are we...

...sts.llvm.org>>; Liu, Yaxun (Sam) <Yaxun.Liu at amd.com<mailto:Yaxun.Liu at amd.com>> Subject: builtins name mangling in SPIR 2.0 Hi all, According to the SPIR 2.0 spec[1], the name of OpenCL builtins are mangled. However, when I compile OpenCl code with Clang 3.9 with the "spir64-unknown-unknown" target, Clang generates IR without mangling the builtins, e.g. for: __kernel void input_zip_int(__global int *in0) { *in0 = get_global_id(0); } clang generates: define spir_kernel void @input_zip_int(i32 addrspace(1)* nocapture %in0) local_unnamed_addr #0 !kernel_arg_addr...