thr3ads.net - llvm dev - [llvm-dev] LIBCLC with LLVM 3.9 Trunk [Apr 2016]

If this information is useful, please help other people find it:
Share via:

mats petersson via llvm-dev

2016-Apr-08 09:06 UTC

[llvm-dev] LIBCLC with LLVM 3.9 Trunk

It's not clear what is actually wrong from your original message, I think
you need to give some more information as to what you are doing: Example
source, what target GPU, compiler error messages or other evidence of
"it's
wrong" (llvm IR, disassembly, etc) ...

--
Mats

On 8 April 2016 at 09:55, Liu Xin via llvm-dev <llvm-dev at
lists.llvm.org>
wrote:
> I built it yesterday. it works. actually, you can only build libclc using
> llvm 3.9.
>
> yes, the interface of writeBitcodeToFile changed. make sure you delete old
> '*.o' files before.
>
> "so the generated GPU kernel code does not have the implementation of
> OpenCl API."
> ~~what's that mean?
>
>
> On Fri, Apr 8, 2016 at 3:19 PM, Heidarshenas, Azin via llvm-dev <
> llvm-dev at lists.llvm.org> wrote:
>
>> Hello there,
>>
>> Has anyone here used the recent LIBCLC (current git version) with the
>> recent git version LLVM (3.9)?
>> It seems to me that the libraries are not compiled properly and so the
>> generated GPU kernel code does not have the implementation of OpenCl
API.
>>
>> Thanks,
>> Azin
>>
>> _______________________________________________
>> LLVM Developers mailing list
>> llvm-dev at lists.llvm.org
>> http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev
>>
>>
>
> _______________________________________________
> LLVM Developers mailing list
> llvm-dev at lists.llvm.org
> http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev
>
>-------------- next part --------------
An HTML attachment was scrubbed...
URL:
<http://lists.llvm.org/pipermail/llvm-dev/attachments/20160408/3d996516/attachment.html>

Heidarshenas, Azin via llvm-dev

2016-Apr-08 16:16 UTC

head link

[llvm-dev] LIBCLC with LLVM 3.9 Trunk

Here is my LLVM version:

LLVM version 3.9.0svn

  Optimized build with assertions.

  Built Apr  5 2016 (15:33:13).

  Default target: x86_64-unknown-linux-gnu

  Host CPU: haswell


I am able to ‘configure' and ‘make' libclc with llvm-config as well but
I am getting a bunch of warnings like following during the make process by
llvm-link:


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/add_sat.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/sub_sat.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/subnormal_helper_func.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/atomic/atomic_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/add_sat_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/clz_if.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/clz_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/sub_sat_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'


WARNING: Linking two modules of different data layouts:
'generic--/lib/subnormal_use_default.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64’


Finally I am able to do ‘make install’ without any errors.


I am following steps in
here<http://stackoverflow.com/questions/8795114/how-to-use-clang-to-compile-opencl-to-ptx-code>
to generate a .bc file for OpenCL kernels; however, after disassembling the
linked bc file I see that only the declarations of OpenCl functions has been
added and no definition is there.

I suspect that libclc libraries are not compiled completely so that the .bc does
not contain any definition from libclc.


Here is .ll version of my linked .bc file (which exactly matches the kernel.ll
before linking with libclc):


; ModuleID = 'kernel.linked.bc'

target datalayout = "e-p:32:32-i64:64-v16:16-v32:32-n16:32:64"

target triple = "nvptx--nvidiacl"


; Function Attrs: noinline nounwind

define void @dataParallel(float addrspace(1)* nocapture readonly %A, float
addrspace(1)* nocapture readonly %B, float addrspace(1)* nocapture %C) #0 {

entry:

  %call = tail call i32 @get_global_id(i32 0) #2

  %mul = shl i32 %call, 2

  %arrayidx = getelementptr inbounds float, float addrspace(1)* %A, i32 %mul

  %0 = load float, float addrspace(1)* %arrayidx, align 4, !tbaa !9

  %arrayidx2 = getelementptr inbounds float, float addrspace(1)* %B, i32 %mul

  %1 = load float, float addrspace(1)* %arrayidx2, align 4, !tbaa !9

  %add3 = fadd float %0, %1

  %arrayidx5 = getelementptr inbounds float, float addrspace(1)* %C, i32 %mul

  store float %add3, float addrspace(1)* %arrayidx5, align 4, !tbaa !9

  %add6 = or i32 %mul, 1

  %arrayidx7 = getelementptr inbounds float, float addrspace(1)* %A, i32 %add6

  %2 = load float, float addrspace(1)* %arrayidx7, align 4, !tbaa !9

  %arrayidx9 = getelementptr inbounds float, float addrspace(1)* %B, i32 %add6

  %3 = load float, float addrspace(1)* %arrayidx9, align 4, !tbaa !9

  %sub = fsub float %2, %3

  %arrayidx11 = getelementptr inbounds float, float addrspace(1)* %C, i32 %add6

  store float %sub, float addrspace(1)* %arrayidx11, align 4, !tbaa !9

  %add12 = or i32 %mul, 2

  %arrayidx13 = getelementptr inbounds float, float addrspace(1)* %A, i32 %add12

  %4 = load float, float addrspace(1)* %arrayidx13, align 4, !tbaa !9

  %arrayidx15 = getelementptr inbounds float, float addrspace(1)* %B, i32 %add12

  %5 = load float, float addrspace(1)* %arrayidx15, align 4, !tbaa !9

  %mul16 = fmul float %4, %5

  %arrayidx18 = getelementptr inbounds float, float addrspace(1)* %C, i32 %add12

  store float %mul16, float addrspace(1)* %arrayidx18, align 4, !tbaa !9

  %add19 = or i32 %mul, 3

  %arrayidx20 = getelementptr inbounds float, float addrspace(1)* %A, i32 %add19

  %6 = load float, float addrspace(1)* %arrayidx20, align 4, !tbaa !9

  %arrayidx22 = getelementptr inbounds float, float addrspace(1)* %B, i32 %add19

  %7 = load float, float addrspace(1)* %arrayidx22, align 4, !tbaa !9

  %div = fdiv float %6, %7, !fpmath !13

  %arrayidx24 = getelementptr inbounds float, float addrspace(1)* %C, i32 %add19

  store float %div, float addrspace(1)* %arrayidx24, align 4, !tbaa !9

  ret void

}


declare i32 @get_global_id(i32) #1


attributes #0 = { noinline nounwind
"disable-tail-calls"="false"
"less-precise-fpmad"="false"
"no-frame-pointer-elim"="true"
"no-frame-pointer-elim-non-leaf"
"no-infs-fp-math"="false"
"no-nans-fp-math"="false"
"stack-protector-buffer-size"="8"
"unsafe-fp-math"="false"
"use-soft-float"="false" }

attributes #1 = { "disable-tail-calls"="false"
"less-precise-fpmad"="false"
"no-frame-pointer-elim"="true"
"no-frame-pointer-elim-non-leaf"
"no-infs-fp-math"="false"
"no-nans-fp-math"="false"
"stack-protector-buffer-size"="8"
"unsafe-fp-math"="false"
"use-soft-float"="false" }

attributes #2 = { nounwind }









From: <mats.o.petersson at googlemail.com<mailto:mats.o.petersson at
googlemail.com>> on behalf of mats petersson <mats at
planetcatfish.com<mailto:mats at planetcatfish.com>>
Date: Friday, April 8, 2016 at 4:06 AM
To: Liu Xin <navy.xliu at gmail.com<mailto:navy.xliu at gmail.com>>
Cc: Azin Heidarshenas <heidars2 at illinois.edu<mailto:heidars2 at
illinois.edu>>, via llvm-dev <llvm-dev at
lists.llvm.org<mailto:llvm-dev at lists.llvm.org>>
Subject: Re: [llvm-dev] LIBCLC with LLVM 3.9 Trunk

It's not clear what is actually wrong from your original message, I think
you need to give some more information as to what you are doing: Example source,
what target GPU, compiler error messages or other evidence of "it's
wrong" (llvm IR, disassembly, etc) ...

--
Mats

On 8 April 2016 at 09:55, Liu Xin via llvm-dev <llvm-dev at
lists.llvm.org<mailto:llvm-dev at lists.llvm.org>> wrote:
I built it yesterday. it works. actually, you can only build libclc using llvm
3.9.

yes, the interface of writeBitcodeToFile changed. make sure you delete old
'*.o' files before.

"so the generated GPU kernel code does not have the implementation of
OpenCl API."
~~what's that mean?


On Fri, Apr 8, 2016 at 3:19 PM, Heidarshenas, Azin via llvm-dev <llvm-dev at
lists.llvm.org<mailto:llvm-dev at lists.llvm.org>> wrote:
Hello there,

Has anyone here used the recent LIBCLC (current git version) with the recent git
version LLVM (3.9)?
It seems to me that the libraries are not compiled properly and so the generated
GPU kernel code does not have the implementation of OpenCl API.

Thanks,
Azin

_______________________________________________
LLVM Developers mailing list
llvm-dev at lists.llvm.org<mailto:llvm-dev at lists.llvm.org>
http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev<https://urldefense.proofpoint.com/v2/url?u=http-3A__lists.llvm.org_cgi-2Dbin_mailman_listinfo_llvm-2Ddev&d=BQMFaQ&c=8hUWFZcy2Z-Za5rBPlktOQ&r=p2pwXNeELP9EUTLxgrxKT2AuaJtLgOAJnwrQ2fUq8LU&m=_pWa85Q85UU31h6Ps2XgXez8ahUNhkraZw7LemHqTfE&s=WJiO7CteVbNhilii4FFLF6KpKbtkaB6dF06vmS5GIwk&e=>



_______________________________________________
LLVM Developers mailing list
llvm-dev at lists.llvm.org<mailto:llvm-dev at lists.llvm.org>
http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev<https://urldefense.proofpoint.com/v2/url?u=http-3A__lists.llvm.org_cgi-2Dbin_mailman_listinfo_llvm-2Ddev&d=BQMFaQ&c=8hUWFZcy2Z-Za5rBPlktOQ&r=p2pwXNeELP9EUTLxgrxKT2AuaJtLgOAJnwrQ2fUq8LU&m=_pWa85Q85UU31h6Ps2XgXez8ahUNhkraZw7LemHqTfE&s=WJiO7CteVbNhilii4FFLF6KpKbtkaB6dF06vmS5GIwk&e=>


-------------- next part --------------
An HTML attachment was scrubbed...
URL:
<http://lists.llvm.org/pipermail/llvm-dev/attachments/20160408/73bd1f5b/attachment.html>

Tom Stellard via llvm-dev

2016-Apr-09 01:34 UTC

head link

[llvm-dev] LIBCLC with LLVM 3.9 Trunk

On Fri, Apr 08, 2016 at 04:16:35PM +0000, Heidarshenas, Azin via llvm-dev
wrote:> Here is my LLVM version:
> 
> LLVM version 3.9.0svn
> 
>   Optimized build with assertions.
> 
>   Built Apr  5 2016 (15:33:13).
> 
>   Default target: x86_64-unknown-linux-gnu
> 
>   Host CPU: haswell
> 
> 
> I am able to ‘configure' and ‘make' libclc with llvm-config as well
but I am getting a bunch of warnings like following during the make process by
llvm-link:
> 
Hi,

What command did you use for linking.  In that stack overflow post I
see:



llvm-link libclc/built_libs/nvptx64--nvidiacl.bc test.ll -o test.linked.bc

Which I'm not sure is correct.  Try passing test.ll to llvm-link before
the bitcode library.

-Tom

> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/add_sat.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/sub_sat.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/subnormal_helper_func.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/atomic/atomic_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/add_sat_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/clz_if.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/clz_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'nvptx--nvidiacl/lib/integer/sub_sat_impl.ll.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64'
> 
> 
> WARNING: Linking two modules of different data layouts:
'generic--/lib/subnormal_use_default.bc' is '' whereas
'llvm-link' is 'e-p:32:32-i64:64-v16:16-v32:32-n16:32:64’
> 
> 
> Finally I am able to do ‘make install’ without any errors.
> 
> 
> I am following steps in
here<http://stackoverflow.com/questions/8795114/how-to-use-clang-to-compile-opencl-to-ptx-code>
to generate a .bc file for OpenCL kernels; however, after disassembling the
linked bc file I see that only the declarations of OpenCl functions has been
added and no definition is there.
> 
> I suspect that libclc libraries are not compiled completely so that the .bc
does not contain any definition from libclc.
> 
> 
> Here is .ll version of my linked .bc file (which exactly matches the
kernel.ll before linking with libclc):
> 
> 
> ; ModuleID = 'kernel.linked.bc'
> 
> target datalayout = "e-p:32:32-i64:64-v16:16-v32:32-n16:32:64"
> 
> target triple = "nvptx--nvidiacl"
> 
> 
> ; Function Attrs: noinline nounwind
> 
> define void @dataParallel(float addrspace(1)* nocapture readonly %A, float
addrspace(1)* nocapture readonly %B, float addrspace(1)* nocapture %C) #0 {
> 
> entry:
> 
>   %call = tail call i32 @get_global_id(i32 0) #2
> 
>   %mul = shl i32 %call, 2
> 
>   %arrayidx = getelementptr inbounds float, float addrspace(1)* %A, i32
%mul
> 
>   %0 = load float, float addrspace(1)* %arrayidx, align 4, !tbaa !9
> 
>   %arrayidx2 = getelementptr inbounds float, float addrspace(1)* %B, i32
%mul
> 
>   %1 = load float, float addrspace(1)* %arrayidx2, align 4, !tbaa !9
> 
>   %add3 = fadd float %0, %1
> 
>   %arrayidx5 = getelementptr inbounds float, float addrspace(1)* %C, i32
%mul
> 
>   store float %add3, float addrspace(1)* %arrayidx5, align 4, !tbaa !9
> 
>   %add6 = or i32 %mul, 1
> 
>   %arrayidx7 = getelementptr inbounds float, float addrspace(1)* %A, i32
%add6
> 
>   %2 = load float, float addrspace(1)* %arrayidx7, align 4, !tbaa !9
> 
>   %arrayidx9 = getelementptr inbounds float, float addrspace(1)* %B, i32
%add6
> 
>   %3 = load float, float addrspace(1)* %arrayidx9, align 4, !tbaa !9
> 
>   %sub = fsub float %2, %3
> 
>   %arrayidx11 = getelementptr inbounds float, float addrspace(1)* %C, i32
%add6
> 
>   store float %sub, float addrspace(1)* %arrayidx11, align 4, !tbaa !9
> 
>   %add12 = or i32 %mul, 2
> 
>   %arrayidx13 = getelementptr inbounds float, float addrspace(1)* %A, i32
%add12
> 
>   %4 = load float, float addrspace(1)* %arrayidx13, align 4, !tbaa !9
> 
>   %arrayidx15 = getelementptr inbounds float, float addrspace(1)* %B, i32
%add12
> 
>   %5 = load float, float addrspace(1)* %arrayidx15, align 4, !tbaa !9
> 
>   %mul16 = fmul float %4, %5
> 
>   %arrayidx18 = getelementptr inbounds float, float addrspace(1)* %C, i32
%add12
> 
>   store float %mul16, float addrspace(1)* %arrayidx18, align 4, !tbaa !9
> 
>   %add19 = or i32 %mul, 3
> 
>   %arrayidx20 = getelementptr inbounds float, float addrspace(1)* %A, i32
%add19
> 
>   %6 = load float, float addrspace(1)* %arrayidx20, align 4, !tbaa !9
> 
>   %arrayidx22 = getelementptr inbounds float, float addrspace(1)* %B, i32
%add19
> 
>   %7 = load float, float addrspace(1)* %arrayidx22, align 4, !tbaa !9
> 
>   %div = fdiv float %6, %7, !fpmath !13
> 
>   %arrayidx24 = getelementptr inbounds float, float addrspace(1)* %C, i32
%add19
> 
>   store float %div, float addrspace(1)* %arrayidx24, align 4, !tbaa !9
> 
>   ret void
> 
> }
> 
> 
> declare i32 @get_global_id(i32) #1
> 
> 
> attributes #0 = { noinline nounwind
"disable-tail-calls"="false"
"less-precise-fpmad"="false"
"no-frame-pointer-elim"="true"
"no-frame-pointer-elim-non-leaf"
"no-infs-fp-math"="false"
"no-nans-fp-math"="false"
"stack-protector-buffer-size"="8"
"unsafe-fp-math"="false"
"use-soft-float"="false" }
> 
> attributes #1 = { "disable-tail-calls"="false"
"less-precise-fpmad"="false"
"no-frame-pointer-elim"="true"
"no-frame-pointer-elim-non-leaf"
"no-infs-fp-math"="false"
"no-nans-fp-math"="false"
"stack-protector-buffer-size"="8"
"unsafe-fp-math"="false"
"use-soft-float"="false" }
> 
> attributes #2 = { nounwind }
> 
> 
> 
> 
> 
> 
> 
> 
> 
> From: <mats.o.petersson at googlemail.com<mailto:mats.o.petersson at
googlemail.com>> on behalf of mats petersson <mats at
planetcatfish.com<mailto:mats at planetcatfish.com>>
> Date: Friday, April 8, 2016 at 4:06 AM
> To: Liu Xin <navy.xliu at gmail.com<mailto:navy.xliu at
gmail.com>>
> Cc: Azin Heidarshenas <heidars2 at illinois.edu<mailto:heidars2 at
illinois.edu>>, via llvm-dev <llvm-dev at
lists.llvm.org<mailto:llvm-dev at lists.llvm.org>>
> Subject: Re: [llvm-dev] LIBCLC with LLVM 3.9 Trunk
> 
> It's not clear what is actually wrong from your original message, I
think you need to give some more information as to what you are doing: Example
source, what target GPU, compiler error messages or other evidence of
"it's wrong" (llvm IR, disassembly, etc) ...
> 
> --
> Mats
> 
> On 8 April 2016 at 09:55, Liu Xin via llvm-dev <llvm-dev at
lists.llvm.org<mailto:llvm-dev at lists.llvm.org>> wrote:
> I built it yesterday. it works. actually, you can only build libclc using
llvm 3.9.
> 
> yes, the interface of writeBitcodeToFile changed. make sure you delete old
'*.o' files before.
> 
> "so the generated GPU kernel code does not have the implementation of
OpenCl API."
> ~~what's that mean?
> 
> 
> On Fri, Apr 8, 2016 at 3:19 PM, Heidarshenas, Azin via llvm-dev
<llvm-dev at lists.llvm.org<mailto:llvm-dev at lists.llvm.org>>
wrote:
> Hello there,
> 
> Has anyone here used the recent LIBCLC (current git version) with the
recent git version LLVM (3.9)?
> It seems to me that the libraries are not compiled properly and so the
generated GPU kernel code does not have the implementation of OpenCl API.
> 
> Thanks,
> Azin
> 
> _______________________________________________
> LLVM Developers mailing list
> llvm-dev at lists.llvm.org<mailto:llvm-dev at lists.llvm.org>
>
http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev<https://urldefense.proofpoint.com/v2/url?u=http-3A__lists.llvm.org_cgi-2Dbin_mailman_listinfo_llvm-2Ddev&d=BQMFaQ&c=8hUWFZcy2Z-Za5rBPlktOQ&r=p2pwXNeELP9EUTLxgrxKT2AuaJtLgOAJnwrQ2fUq8LU&m=_pWa85Q85UU31h6Ps2XgXez8ahUNhkraZw7LemHqTfE&s=WJiO7CteVbNhilii4FFLF6KpKbtkaB6dF06vmS5GIwk&e=>
> 
> 
> 
> _______________________________________________
> LLVM Developers mailing list
> llvm-dev at lists.llvm.org<mailto:llvm-dev at lists.llvm.org>
>
http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev<https://urldefense.proofpoint.com/v2/url?u=http-3A__lists.llvm.org_cgi-2Dbin_mailman_listinfo_llvm-2Ddev&d=BQMFaQ&c=8hUWFZcy2Z-Za5rBPlktOQ&r=p2pwXNeELP9EUTLxgrxKT2AuaJtLgOAJnwrQ2fUq8LU&m=_pWa85Q85UU31h6Ps2XgXez8ahUNhkraZw7LemHqTfE&s=WJiO7CteVbNhilii4FFLF6KpKbtkaB6dF06vmS5GIwk&e=>
> 
> 
> _______________________________________________
> LLVM Developers mailing list
> llvm-dev at lists.llvm.org
> http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev

llvm dev - Apr 2016 - LIBCLC with LLVM 3.9 Trunk

[llvm-dev] LIBCLC with LLVM 3.9 Trunk

[llvm-dev] LIBCLC with LLVM 3.9 Trunk

[llvm-dev] LIBCLC with LLVM 3.9 Trunk