[llvm-dev] LIBCLC with LLVM 3.9 Trunk
Tom Stellard via llvm-dev
llvm-dev at lists.llvm.org
Fri Apr 8 18:34:54 PDT 2016
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
More information about the llvm-dev
mailing list