1*9880d681SAndroid Build Coastguard Worker; RUN: llc < %s -mtriple=nvptx64-nvidia-cuda | FileCheck %s 2*9880d681SAndroid Build Coastguard Worker 3*9880d681SAndroid Build Coastguard Worker; // Bitcode int this test case is reduced version of compiled code below: 4*9880d681SAndroid Build Coastguard Worker;extern "C" { 5*9880d681SAndroid Build Coastguard Worker;#line 1 "/source/dir/foo.h" 6*9880d681SAndroid Build Coastguard Worker;__device__ void foo() {} 7*9880d681SAndroid Build Coastguard Worker;#line 2 "/source/dir/bar.cu" 8*9880d681SAndroid Build Coastguard Worker;__device__ void bar() {} 9*9880d681SAndroid Build Coastguard Worker;} 10*9880d681SAndroid Build Coastguard Worker 11*9880d681SAndroid Build Coastguard Worker; CHECK: .file 1 "/source/dir{{/|\\\\}}bar.cu" 12*9880d681SAndroid Build Coastguard Worker; CHECK: .file 2 "/source/dir{{/|\\\\}}foo.h" 13*9880d681SAndroid Build Coastguard Worker 14*9880d681SAndroid Build Coastguard Worker; CHECK-LABEL: @foo 15*9880d681SAndroid Build Coastguard Workerdefine void @foo() !dbg !4 { 16*9880d681SAndroid Build Coastguard Workerbb: 17*9880d681SAndroid Build Coastguard Worker ret void, !dbg !10 18*9880d681SAndroid Build Coastguard Worker} 19*9880d681SAndroid Build Coastguard Worker; CHECK: .loc 2 1 20*9880d681SAndroid Build Coastguard Worker; CHECK: ret 21*9880d681SAndroid Build Coastguard Worker 22*9880d681SAndroid Build Coastguard Worker; CHECK-LABEL: @bar 23*9880d681SAndroid Build Coastguard Workerdefine void @bar() !dbg !7 { 24*9880d681SAndroid Build Coastguard Workerbb: 25*9880d681SAndroid Build Coastguard Worker ret void, !dbg !11 26*9880d681SAndroid Build Coastguard Worker} 27*9880d681SAndroid Build Coastguard Worker; CHECK: .loc 1 2 28*9880d681SAndroid Build Coastguard Worker; CHECK: ret 29*9880d681SAndroid Build Coastguard Worker 30*9880d681SAndroid Build Coastguard Worker!llvm.dbg.cu = !{!0} 31*9880d681SAndroid Build Coastguard Worker!llvm.module.flags = !{!8, !9} 32*9880d681SAndroid Build Coastguard Worker 33*9880d681SAndroid Build Coastguard Worker!0 = distinct !DICompileUnit(language: DW_LANG_C_plus_plus, file: !1, producer: "", isOptimized: false, runtimeVersion: 0, emissionKind: LineTablesOnly, enums: !2) 34*9880d681SAndroid Build Coastguard Worker!1 = !DIFile(filename: "bar.cu", directory: "/source/dir") 35*9880d681SAndroid Build Coastguard Worker!2 = !{} 36*9880d681SAndroid Build Coastguard Worker!4 = distinct !DISubprogram(name: "foo", scope: !5, file: !5, line: 1, type: !6, isLocal: false, isDefinition: true, scopeLine: 1, flags: DIFlagPrototyped, isOptimized: false, unit: !0, variables: !2) 37*9880d681SAndroid Build Coastguard Worker!5 = !DIFile(filename: "foo.h", directory: "/source/dir") 38*9880d681SAndroid Build Coastguard Worker!6 = !DISubroutineType(types: !2) 39*9880d681SAndroid Build Coastguard Worker!7 = distinct !DISubprogram(name: "bar", scope: !1, file: !1, line: 2, type: !6, isLocal: false, isDefinition: true, scopeLine: 2, flags: DIFlagPrototyped, isOptimized: false, unit: !0, variables: !2) 40*9880d681SAndroid Build Coastguard Worker!8 = !{i32 2, !"Dwarf Version", i32 4} 41*9880d681SAndroid Build Coastguard Worker!9 = !{i32 2, !"Debug Info Version", i32 3} 42*9880d681SAndroid Build Coastguard Worker!10 = !DILocation(line: 1, column: 31, scope: !4) 43*9880d681SAndroid Build Coastguard Worker!11 = !DILocation(line: 2, column: 31, scope: !7) 44