[BUG] Kernel fixup does not adjust debug metadata arguments

[gmarkall]

The kernel fixup added in #50 removes the return value placeholder from the argument list, but does not remove it from the debug metadata. Thus, a kernel like:

@cuda.jit("void(int32, int32)", debug=True, opt=False)
def k(x, y): 
    x + y

will have the following prototype in LLVM IR:

define void @"k"(i32 %"arg.x", i32 %"arg.y") !dbg !10

(function name demangled for clarity)

but the debug metadata is:

!3 = !DIBasicType(encoding: DW_ATE_unsigned, name: "i8", size: 8)
!4 = !DIDerivedType(baseType: !3, size: 64, tag: DW_TAG_pointer_type)
!5 = !DIDerivedType(baseType: !4, size: 64, tag: DW_TAG_pointer_type)
!6 = !DIBasicType(encoding: DW_ATE_signed, name: "int32", size: 32)
!7 = !DIBasicType(encoding: DW_ATE_signed, name: "int32", size: 32)
!8 = !{ !5, !6, !7 }
!9 = !DISubroutineType(types: !8)
!10 = distinct !DISubprogram(file: !1, isDefinition: true, isLocal: false, isOptimized: true,
                             line: 5, linkageName: "k", name: "k", scope: !1, scopeLine: 5, type: !9, unit: !2)

The issue being that the arguments in !8 are !{ !5, !6, !7 }. !5 is the placeholder for a return value (the equivalent of void**), which should have been removed to leave only !{ !6, !7 }.