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