[llvm-dev] builtins name mangling in SPIR 2.0

Mon Sep 19 11:59:59 PDT 2016

On Mon, Sep 19, 2016 at 3:53 AM, Bader, Alexey <alexey.bader at intel.com>
wrote:

> I’d like to clarify that old SPIR 1.2 and SPIR 2.0 formats are flavors of
> LLVM IR version 3.2/3.4, so clang 3.9 will never be able to produce
> “conformant” SPIR programs. Considering this I think it makes sense to
> evaluate the benefit of features from SPIR 1.2/2.0 specification we are
> trying to implement in ToT clang.
>
>
>
> Could you clarify the problem you are trying to solve by mangling printf
> function?
>
I am not solving any problem by mangling the printf. I just confused when I
saw the header and the SPIR 2.0 "Provisional" specs do not agree :)

>
>
> Thanks,
>
> Alexey
>
>
>
> *From:* Anastasia Stulova [mailto:Anastasia.Stulova at arm.com]
> *Sent:* Sunday, September 18, 2016 10:59 PM
> *To:* Hongbin Zheng <etherzhhb at gmail.com>; Liu, Yaxun (Sam) <
> Yaxun.Liu at amd.com>
> *Cc:* cfe-dev at lists.llvm.org; llvm-dev <llvm-dev at lists.llvm.org>; Bader,
> Alexey <alexey.bader at intel.com>; nd <nd at arm.com>
>
> *Subject:* Re: builtins name mangling in SPIR 2.0
>
>
>
>
> I don't see any problem mangling it to be honest even though there seems
> to be only one prototype anyways.
>
> We could add restrict in as well.
>
> Cheers,
> Anastasia
> ------------------------------
>
> *From:* Hongbin Zheng <etherzhhb at gmail.com>
> *Sent:* 17 September 2016 05:32:54
> *To:* Liu, Yaxun (Sam)
> *Cc:* cfe-dev at lists.llvm.org; llvm-dev; Bader, Alexey (
> alexey.bader at intel.com); Anastasia Stulova
> *Subject:* Re: builtins name mangling in SPIR 2.0
>
>
>
>
>
>
>
> On Fri, Sep 16, 2016 at 8:10 AM, Liu, Yaxun (Sam) <Yaxun.Liu at amd.com>
> wrote:
>
> + Alexey Anastasia
>
>
>
> According to SPIR spec v1.2 s2.10.3
>
>
>
> *2.10.3 The printf function*
> The printf function is supported, and is mangled according to its
> prototype as follows:
> int printf(constant char * restrict fmt, ... )
>
> clang 3.9 will drop the 'restrict' when it mangle this name, or I am
> missing some commandline argument?
>
>
>
> Thanks
>
> Hongbin
>
>
>
> Note that the ellipsis formal argument (...) is mangled to argument type
> specifier z
>
>
>
> It seems printf should be mangled.
>
>
>
> Alexey/Anastasia,
>
>
>
> What do you think? Thanks.
>
>
>
> Sam
>
>
>
> *From:* Hongbin Zheng [mailto:etherzhhb at gmail.com]
> *Sent:* Friday, September 16, 2016 1:54 AM
> *To:* Liu, Yaxun (Sam) <Yaxun.Liu at amd.com>
> *Cc:* cfe-dev at lists.llvm.org; llvm-dev <llvm-dev at lists.llvm.org>
> *Subject:* Re: builtins name mangling in SPIR 2.0
>
>
>
> Hi Sam and others,
>
>
>
> I saw that in [1], printf is mangled to _Z6printfPrU3AS2cz, while in
> clang's opencl-c.h[2], printf does not have the overload attribute:
>
> int printf(__constant const char* st, ...); (and it is different from the
> standard, which is printf(restrict __constant char *, ...))
>
> I try the following code:
>
> #include <opencl-c.h>
>
>
>
> __kernel void vadd(__global const int* a, __global const int* b, __global
> int* c) {
>
>   printf("aaaaa");
>
> }
>
>
>
> and get a printf that is not mangled:
>
>   %call = tail call spir_func i32 (i8 addrspace(2)*, ...) @printf(i8
> addrspace(2)* getelementptr inbounds ([6 x i8], [6 x i8] addrspace(2)*
> @.str, i64 0, i64 0))
>
>
>
> with the following command line:
>
> clang -cc1 -internal-isystem /wrk/xsjhdnobkup2/hongbinz/
> omp/build-llvm/bin/../lib/clang/3.9.0/include -nostdsysteminc -S
> -emit-llvm -o -
>
>
>
> Is this the correct behavior?
>
>
>
> Thanks
>
> Hongbin
>
>
>
> [1] https://github.com/KhronosGroup/SPIR-Tools/wiki/
> SPIR-2.0-built-in-functions
>
> [2] http://clang.llvm.org/doxygen/opencl-c_8h_source.html
>
>
>
> On Mon, Sep 12, 2016 at 1:43 PM, Hongbin Zheng <etherzhhb at gmail.com>
> wrote:
>
> Thanks a lot.
>
>
>
> On Mon, Sep 12, 2016 at 1:42 PM, Liu, Yaxun (Sam) <Yaxun.Liu at amd.com>
> wrote:
>
> If you use the default header file under clang/lib/Headers/opencl-c.h,
> get_global_id will be mangled.
>
>
>
> If you want to declare get_global_id in your own header, add
> __attribute__((overloadable)), then it will be mangled.
>
>
>
> Sam
>
>
>
> *From:* Hongbin Zheng [mailto:etherzhhb at gmail.com]
> *Sent:* Monday, September 12, 2016 4:21 PM
> *To:* cfe-dev at lists.llvm.org; llvm-dev <llvm-dev at lists.llvm.org>; Liu,
> Yaxun (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 @input_zip_int(i32 addrspace(1)* nocapture %in0)
> local_unnamed_addr #0 !kernel_arg_addr_space !3 !kernel_arg_access_qual !4
> !kernel_arg_type !5 !kernel_arg_base_type !5 !kernel_arg_type_qual !6 {
>
> entry:
>
>   %call = tail call spir_func i128 @get_global_id(i32 0) #2
>
>   %conv = trunc i128 %call to i32
>
>   store i32 %conv, i32 addrspace(1)* %in0, align 4, !tbaa !7
>
>   ret void
>
> }
>
>
>
> In this case, get_global_id is not mangled to _Z13get_global_idj according
> to [2].
>
>
>
> Is there an option for clang or an LLVM clang to do the mangling for spir
> builtins?
>
>
>
> Thanks
>
> Hongbin
>
>
>
>
>
> [1] https://www.khronos.org/registry/spir/specs/spir_spec-2.0.pdf, page 36
>
> [2] https://github.com/KhronosGroup/SPIR-Tools/wiki/
> SPIR-2.0-built-in-functions
>
>
>
>
>
>
>
>
> --------------------------------------------------------------------
> Joint Stock Company Intel A/O
> Registered legal address: Krylatsky Hills Business Park,
> 17 Krylatskaya Str., Bldg 4, Moscow 121614,
> Russian Federation
>
> This e-mail and any attachments may contain confidential material for
> the sole use of the intended recipient(s). Any review or distribution
> by others is strictly prohibited. If you are not the intended
> recipient, please contact the sender and delete all copies.
>
-------------- next part --------------
An HTML attachment was scrubbed...
URL: <http://lists.llvm.org/pipermail/llvm-dev/attachments/20160919/aaac288b/attachment.html>