similar to: [LLVMdev] NVPTX: __iAtomicCAS support ?

> -----Original Message----- > From: Dmitry N. Mikushin [mailto:maemarcus at gmail.com] > Sent: Wednesday, May 16, 2012 5:44 AM > To: LLVM-Dev > Cc: Justin Holewinski > Subject: NVPTX: __iAtomicCAS support ? > > Dear colleagues, > > I'm looking if we can replace nvopencc with LLVM NVPTX in our project. > It turns NVPTX won't work with the code nvopencc

Dear LLVM, I'm trying to understand why the attached IR code works for x86_64 target and fails for nvptx64, because of unimplemented expand during the target lowering. Any ideas? Just change the target triple to x86_64-unknown-unknown, and the same IR code could we successfully codegen-ed for x86_64. Thanks, - Dima. dmikushin at dmikushin-desktop:~/Desktop$ gdb ~/sandbox/bin/llc GNU gdb

Hi again, Kind people on #llvm helped me to utilize bugpoint to reduce the previously submitted test case. For record, it code be done with the following command: $ bugpoint -llc-safe test.ll The resulting IR is attached, and it is crashing in the same way. Is it a valid code? dmikushin at hp2:~/forge/kernelgen/branches/tests_lnt/behavior/sincos> llc test.ll.1 This action is not supported

Dear Zhang, Compiler ends up invoking cc1 (the backend) anyways. So if you would like to invoke it by hand, the only thing to know is the right combination of options. Try to use the compiler verbose option "-v". It will show you how exactly clang invokes the backend: > clang -v -c showdebug.c clang version 3.2 (trunk 156703) Target: x86_64-unknown-linux-gnu Thread model: posix

Hi, After building out project in release mode, caught an assertion, which we have not seen before: hello_f: /tmp/rpmbuild_debug/BUILD/llvm/build/include/llvm/ADT/DenseMap.h:126: void llvm::DenseMap<KeyT, ValueT, KeyInfoT>::clear() [with KeyT = llvm::MachineBasicBlock*, ValueT = <unnamed>::BlockChain*, KeyInfoT = llvm::DenseMapInfo<llvm::MachineBasicBlock*>]: Assertion

On Thu, Aug 2, 2012 at 8:56 AM, Xinglin Zhang <xinglinzh at gmail.com> wrote: > Hi, > > I am quite new to LLVM. I just compiled LLVM and clang on Ubuntu11.10 then > followed the tutorial http://llvm.org/docs/DebuggingJITedCode.html > > clang -cc1 -O0 -g -emit-llvm showdebug.c > > > where showdebug.c contains: > > #include<stdio.h> > int main() >

Dear NVPTX community, I've create a bug http://llvm.org/bugs/show_bug.cgi?id=13521 with reprocase for this issue. Please, help us to fix it. Last 1,5 months we regularly encounter & workaround or fix 1-2 bugs per week in NVPTX backend. This is definitely not the amount of work we can completely serve ourselves... We would really really appreciate some collaboration. Thanks, - D.

Hi, I am quite new to LLVM. I just compiled LLVM and clang on Ubuntu11.10 then followed the tutorial http://llvm.org/docs/DebuggingJITedCode.html clang -cc1 -O0 -g -emit-llvm showdebug.c where showdebug.c contains: #include<stdio.h> int main() { printf("hello\n"); return 0; } But I got Fatal error: 'stdio.h' file not found. However, clang showdebug.c has no

Unfortunately, I cannot reproduce this. Based on your bugzilla comment, it does look like a mis-compile with your system compiler. Does the same issue occur if you build LLVM as static libraries? On 08/03/2012 12:24 AM, Dmitry N. Mikushin wrote: > Dear NVPTX community, > > I've create a bug http://llvm.org/bugs/show_bug.cgi?id=13521 with > reprocase for this issue. > >

Dear NVPTX community, PTXAS fails to compile the ptx code generated by NVPTX. Is it an issue of backend or an issue of PTXAS or a known reasonable restriction? Thanks, - Dima. > cat test.ll ; ModuleID = '__kernelgen_main_module' target datalayout = "e-p:64:64-i64:64:64-f64:64:64-n1:8:16:32:64" target triple = "ptx64-unknown-unknown" %struct.__st_parameter_dt.0.4

In ptx, variables need to be defined before referenced. NVPTX emits the global variables in the order as in the LLVM IR and does not sort them. It is a bug in the NVPTX backend. Thanks. Yuan From: Dmitry N. Mikushin [mailto:maemarcus at gmail.com] Sent: Wednesday, July 18, 2012 7:44 AM To: LLVM-Dev Cc: Justin Holewinski; Yuan Lin Subject: [NVPTX] PTXAS - Unimplemented feature: labels as

Dear all, Looks like the NVPTX backend cannot handle array-of-arrays contant (please see the reporocase below). Is it supposed to work? Any ideas how to get it working? Important for our target applications. Thanks, - Dima. $ cat test.ll ; ModuleID = '__kernelgen_main_module' target datalayout =

Hi all, considering following .ll file ; ModuleID = 'vect3x.ll' target triple = "armv7-none-linux-gnueabi" define arm_aapcscc void @test_hi_char8(i8* %.T0351, <8 x i8>* nocapture %srcA, <4 x i8>* nocapture %dst) noinline { L.entry: %0 = tail call arm_aapcscc i32 (...)* @get_global_id(i8* %.T0351, i32 0) %1 = bitcast <8 x i8>* %srcA to <4 x i8>*

Sounds like a bug in vector promote. If I restore this flag and use -promote-elements=0 everything works for me. Please fill a PR in LLVM bugzilla and assign to Nadav. On Mon, Jun 25, 2012 at 5:04 PM, Sebastien DELDON-GNB <sebastien.deldon at st.com> wrote: > Hi all, > > > considering following .ll file > > ; ModuleID = 'vect3x.ll' > target triple =

NVCC successfully handles the same IR, if we try to process the same .cu file with clang+nvptx and nvcc: CLANG/NVPTX: ============= $ cat dayofweek.cu __attribute__((device)) char yweek[7][4] = { "MON", "TUE", "WED", "THU", "FRI", "SAT", "SUN" }; $ clang -cc1 -emit-llvm -fcuda-is-device dayofweek.cu -o dayofweek.ll $ cat

I think our test case demonstrates that requiring the array item being initialized to be constant is incorrect. NVPTX does not crash anymore and produces correct result with the following change: --- NVPTXAsmPrinter.cpp 2012-09-03 15:14:00.000000000 +0200 +++ NVPTXAsmPrinter.cpp 2012-09-04 15:47:17.859398193 +0200 @@ -1890,17 +1890,15 @@ case Type::ArrayTyID: case Type::VectorTyID: case

On 09/04/2012 09:57 AM, Dmitry N. Mikushin wrote: > I think our test case demonstrates that requiring the array item being > initialized to be constant is incorrect. NVPTX does not crash anymore > and produces correct result with the following change: > > --- NVPTXAsmPrinter.cpp 2012-09-03 15:14:00.000000000 +0200 > +++ NVPTXAsmPrinter.cpp 2012-09-04 15:47:17.859398193 +0200 >

Hello, FYI, this is a bug http://llvm.org/bugs/show_bug.cgi?id=13324 When compiling the following code for sm_20, func params are by some reason given with .align 0, which is invalid. Problem does not occur if compiled for sm_10. > cat test.ll ; ModuleID = '__kernelgen_main_module' target datalayout = "e-p:64:64-i64:64:64-f64:64:64-n1:8:16:32:64" target triple =

Hi All, I tried build a version of LLVM pulled from the git repo on my raspberry pi. I got close to the finish line then: llvm[2]: Building Debug+Asserts Archive Library libLLVMDebugInfo.a make[2]: Leaving directory `/home/pi/llvm-git/lib/DebugInfo' make[1]: Leaving directory `/home/pi/llvm-git/lib' make[1]: Entering directory `/home/pi/llvm-git/tools/llvm-shlib' llvm[1]: Linking

On 01.09.2012, at 15:02, Keith Sheppard <keithshep at gmail.com> wrote: > Hi All, > > I tried build a version of LLVM pulled from the git repo on my > raspberry pi. I got close to the finish line then: > > llvm[2]: Building Debug+Asserts Archive Library libLLVMDebugInfo.a > make[2]: Leaving directory `/home/pi/llvm-git/lib/DebugInfo' > make[1]: Leaving