#0 0x00007ffff5ba19ce in llvm::TargetRegisterInfo::getRegClass (this=0xe52b50, i=4294967295) at /work_local/llvm/include/llvm/Target/TargetRegisterInfo.h:561 No locals. #1 0x00007ffff5ba15e6 in (anonymous namespace)::AMDGPUDAGToDAGISel::PostprocessISelDAG (this=0x704760) at /work_local/llvm/lib/Target/R600/AMDILISelDAGToDAG.cpp:680 Val = {Node = 0xe245b0, ResNo = 0} TRI = 0xe52b50 Reg = 2147483653 RC = 0x7ffff5e57b80 Desc = {Opcode = 12, NumOperands = 1, NumDefs = 1, SchedClass = 0, Size = 0, Flags = 33554437, TSFlags = 0, ImplicitUses = 0x0, ImplicitDefs = 0x0, OpInfo = 0x7ffff58ef000} Node = 0xe28b10 MachineNode = 0xe25ed0 ResNode = 0xe247b0 I = {> = {}, NodePtr = 0xe28b10} E = {> = {}, NodePtr = 0x704a58} Lowering = @0x6d8270: { = { = {_vptr.TargetLoweringBase = 0x7ffff5e595b0, TM = @0x99c2e0, TD = 0x99c430, TLOF = @0xe4f910, PointerTy = {SimpleTy = llvm::MVT::i64}, IsLittleEndian = true, SelectIsExpensive = true, IntDivIsCheap = false, BypassSlowDivWidths = { >, unsigned int, unsigned int, llvm::DenseMapInfo >> = {}, Buckets = 0x0, NumEntries = 0, NumTombstones = 0, NumBuckets = 0}, Pow2DivIsCheap = false, JumpIsExpensive = true, UseUnderscoreSetJmp = false, UseUnderscoreLongJmp = false, SupportJumpTables = true, MinimumJumpTableEntries = 4, BooleanContents = llvm::TargetLoweringBase::UndefinedBooleanContent, BooleanVectorContents = llvm::TargetLoweringBase::UndefinedBooleanContent, SchedPreferenceInfo = llvm::Sched::RegPressure, JumpBufSize = 0, JumpBufAlignment = 0, MinStackArgumentAlignment = 1, MinFunctionAlignment = 0, PrefFunctionAlignment = 0, PrefLoopAlignment = 0, InsertFencesForAtomic = false, StackPointerRegisterToSaveRestore = 0, ExceptionPointerRegister = 0, ExceptionSelectorRegister = 0, RegClassForVT = {0x0, 0x7ffff5e57b80, 0x0, 0x0, 0x7ffff5e573c0, 0x7ffff5e57b80, 0x7ffff5e57d80, 0x0, 0x7ffff5e573c0, 0x0 , 0x7ffff5e57d80, 0x7ffff5e57e00, 0x7ffff5e57e80, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x7ffff5e573c0, 0x7ffff5e57b40, 0x7ffff5e57d00, 0x7ffff5e57dc0, 0x7ffff5e57e40, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x7ffff5e57b40, 0x7ffff5e57d00, 0x7ffff5e57dc0, 0x7ffff5e57e40, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0}, NumRegistersForVT = '\001' , "\002", '\001' , " \001\001\001\001\001\001\002\004\b\020\002\001\001\001\001\002\004\b\001\001\000\001", RegisterTypeForVT = {{SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::i1}, {SimpleTy = llvm::MVT::i32}, { SimpleTy = llvm::MVT::i32}, {SimpleTy = llvm::MVT::i32}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::i128}, {SimpleTy = llvm::MVT::f16}, {SimpleTy = llvm::MVT::f32}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::f80}, {SimpleTy = llvm::MVT::i128}, {SimpleTy = llvm::MVT::f64}, { SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, {SimpleTy = llvm::MVT::v8i32}, {SimpleTy = llvm::MVT::v16i8}, {SimpleTy = llvm::MVT::v32i8}, {SimpleTy = llvm::MVT::v64i8}, {SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, {SimpleTy = llvm::MVT::v8i32}, { SimpleTy = llvm::MVT::v16i8}, {SimpleTy = llvm::MVT::v32i8}, {SimpleTy = llvm::MVT::v64i8}, {SimpleTy = llvm::MVT::i32}, {SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, {SimpleTy = llvm::MVT::v8i32}, {SimpleTy = llvm::MVT::v16i32}, {SimpleTy = llvm::MVT::i32}, { SimpleTy = llvm::MVT::v1i32}, {SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, {SimpleTy = llvm::MVT::v8i32}, {SimpleTy = llvm::MVT::v16i32}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::i64}, { SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::f16}, {SimpleTy = llvm::MVT::v2f32}, {SimpleTy = llvm::MVT::v4f32}, {SimpleTy = llvm::MVT::v8f32}, {SimpleTy = llvm::MVT::v16f32}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::i64}, { SimpleTy = llvm::MVT::x86mmx}, {SimpleTy = llvm::MVT::Glue}, {SimpleTy = llvm::MVT::isVoid}, {SimpleTy = llvm::MVT::Untyped}}, RepRegClassForVT = {0x0, 0x7ffff5e57e80, 0x0, 0x0, 0x7ffff5e57e40, 0x7ffff5e57e80, 0x7ffff5e57e80, 0x0, 0x7ffff5e57e40, 0x0 , 0x7ffff5e57e80, 0x7ffff5e57e80, 0x7ffff5e57e80, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x7ffff5e57e40, 0x7ffff5e57e40, 0x7ffff5e57e40, 0x7ffff5e57e40, 0x7ffff5e57e40, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x7ffff5e57e40, 0x7ffff5e57e40, 0x7ffff5e57e40, 0x7ffff5e57e40, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0}, RepRegClassCostForVT = "\000\001\000\000\001\001\001\000\001", '\000' , "\001\001\001\000\000\000\000\000\000\001\001\001\001\001\000\000\000\000\000\000\001\001\001\001\000\000\000\000\000\000", TransformToType = {{SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::i1}, { SimpleTy = llvm::MVT::i32}, {SimpleTy = llvm::MVT::i32}, {SimpleTy = llvm::MVT::i32}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::i128}, {SimpleTy = llvm::MVT::f16}, {SimpleTy = llvm::MVT::f32}, {SimpleTy = llvm::MVT::i64}, {SimpleTy = llvm::MVT::f80}, {SimpleTy = llvm::MVT::i128}, { SimpleTy = llvm::MVT::f64}, {SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, {SimpleTy = llvm::MVT::v8i32}, {SimpleTy = llvm::MVT::v16i8}, {SimpleTy = llvm::MVT::v32i8}, {SimpleTy = llvm::MVT::v64i8}, {SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, { SimpleTy = llvm::MVT::v8i32}, {SimpleTy = llvm::MVT::v16i8}, {SimpleTy = llvm::MVT::v32i8}, {SimpleTy = llvm::MVT::v64i8}, {SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, {SimpleTy = llvm::MVT::v8i32}, {SimpleTy = llvm::MVT::v16i32}, { SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::v1i32}, {SimpleTy = llvm::MVT::v2i32}, {SimpleTy = llvm::MVT::v4i32}, {SimpleTy = llvm::MVT::v8i32}, {SimpleTy = llvm::MVT::v16i32}, {SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::Other}, { SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::v2f32}, {SimpleTy = llvm::MVT::v4f32}, {SimpleTy = llvm::MVT::v8f32}, {SimpleTy = llvm::MVT::v16f32}, {SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::Other}, { SimpleTy = llvm::MVT::Other}, {SimpleTy = llvm::MVT::x86mmx}, {SimpleTy = llvm::MVT::Glue}, {SimpleTy = llvm::MVT::isVoid}, {SimpleTy = llvm::MVT::Untyped}}, OpActions = { '\000' , "\003", '\000' , "\002\002\002", '\000' , "\002\000\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\002", '\000' , "\002\002\003", '\000' , "\002\002\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , "\003", '\000' , '\000' , "\003\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002", '\000' , "\002\002\002\002", '\000' , "\003", '\000' , "\002\002\003", '\000' , '\000' , "\003\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002", '\000' , "\002\002\002\002", '\000' , "\003", '\000' , "\002\002\003", '\000' , '\000' , "\003\002\002\002\002\002\002\003\000\002\002\002\002", '\000' , "\002\000\000\000\002", '\000' , "\002\000\002\002\002\002\000\000\000\000\003\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\002\000\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\000\000\002\002\000\000\000\000\000\000\000\000\002\002\002\002\000\000\000\000\000\000\000\000\000\003\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002", '\000' , "\002\000\000\000\002", '\000' , "\002\002\002\002\002\002\002\002\002\002", '\000' , "\000\000\002\002\002\002\000\000\000\000\002\000\002\002", '\000' , "\003\000\002\000\002\002\000\000\000\002\002\002\002", '\000' , "\002\000\000\000\002", '\000' , "\003\000\000\000\000\000\000\000\000\000\000\003\000\000\000\000\002", '\000' , "\002\000\002\002\000\000\002\000\002\000\000\001\001\000\000\002\002\003", '\000' , "\000\000\002\002\002\002\000\000\000\000\002\002\002\002", '\000' , "\003\000\002\000\002\002\000\000\000\002\002\002\002", '\000' , "\002\000\000\000\002", '\000' , "\003\000\000\000\000\002", '\000' , "\002\002\002\002\002\002\002\002\002\002\000\000\000\000\000\002\002\003", '\000' , '\000' , "\002", '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002", '\000' , "\002\000\000\000\002", '\000' , "\002\002\002\002\002\002\002\002\002\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\003\002\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\003\002\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\003\002\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\003\002\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\003\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\003\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\002\000\000\000\002\000\000\002", '\000' , '\000' , "\002\000\000\000\002\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\002\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002\000\002\002", '\000' , "\002\000\000\000\000\002\002\002\002\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\003\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\003\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\001\001\000\000\002\002\003", '\000' , '\000' , "\002\000\000\000\002\000\000\002", '\000' , '\000' , "\002\000\000\000\002\000\000\002", '\000' , '\000' , "\003\000\002\000\002\002\002\000\000\002\002\002\002", '\000' , "\002\000\000\000\002\000\000\002", '\000' , "\002\000\000\000\000\000\000\000\000\000\000\003", '\000' , "\002\002\003", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' , '\000' , "\002\000\000\000\002", '\000' }, LoadExtActions = {"\000\000\000" }, TruncStoreActions = {'\000' }, IndexedModeActions = {"\000\"\"\"\"" }, CondCodeActions = {{0, 0} }, ValueTypeActions = { ValueTypeActions = "\000\000\001\001\000\000\000\000\000\003\000\003\004\001\001\001\001\001\001\001\001\001\000\000\000\005\001\001\001\001\006\000\000\000\000\000\005\006\006\006\006\006\000\000\000\000\006\006\006\000\000\000"}, AvailableRegClasses = {, std::allocator > >> = { _M_impl = { >> = {<__gnu_cxx::new_allocator >> = {}, }, _M_start = 0xe52c90, _M_finish = 0xe52da0, _M_end_of_storage = 0xe52e90}}, }, TargetDAGCombineArray = '\000' , "\003\000\000\000\000\000\000\000\000\000", PromoteToType = {_M_t = { _M_impl = { const, llvm::MVT::SimpleValueType> > >> = {<__gnu_cxx::new_allocator const, llvm::MVT::SimpleValueType> > >> = {}, }, _M_key_compare = {, std::pair, bool>> = {}, }, _M_header = {_M_color = std::_S_red, _M_parent = 0xe4bc10, _M_left = 0xe50ce0, _M_right = 0xe4c180}, _M_node_count = 4}}}, LibcallRoutineNames = {0x7ffff183eb90 "__ashlhi3", 0x7ffff183eb9a "__ashlsi3", 0x7ffff183eba4 "__ashldi3", 0x7ffff183ebae "__ashlti3", 0x7ffff183ebb8 "__lshrhi3", 0x7ffff183ebc2 "__lshrsi3", 0x7ffff183ebcc "__lshrdi3", 0x7ffff183ebd6 "__lshrti3", 0x7ffff183ebe0 "__ashrhi3", 0x7ffff183ebea "__ashrsi3", 0x7ffff183ebf4 "__ashrdi3", 0x7ffff183ebfe "__ashrti3", 0x7ffff183ec08 "__mulqi3", 0x7ffff183ec11 "__mulhi3", 0x7ffff183ec1a "__mulsi3", 0x7ffff183ec23 "__muldi3", 0x7ffff183ec2c "__multi3", 0x7ffff183ec35 "__mulosi4", 0x7ffff183ec3f "__mulodi4", 0x7ffff183ec49 "__muloti4", 0x7ffff183ec53 "__divqi3", 0x7ffff183ec5c "__divhi3", 0x7ffff183ec65 "__divsi3", 0x7ffff183ec6e "__divdi3", 0x7ffff183ec77 "__divti3", 0x7ffff183ec80 "__udivqi3", 0x7ffff183ec8a "__udivhi3", 0x7ffff183ec94 "__udivsi3", 0x7ffff183ec9e "__udivdi3", 0x7ffff183eca8 "__udivti3", 0x7ffff183ecb2 "__modqi3", 0x7ffff183ecbb "__modhi3", 0x7ffff183ecc4 "__modsi3", 0x7ffff183eccd "__moddi3", 0x7ffff183ecd6 "__modti3", 0x7ffff183ecdf "__umodqi3", 0x7ffff183ece9 "__umodhi3", 0x7ffff183ecf3 "__umodsi3", 0x7ffff183ecfd "__umoddi3", 0x7ffff183ed07 "__umodti3", 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x0, 0x7ffff183ed11 "__negsi2", 0x7ffff183ed1a "__negdi2", 0x7ffff183ed23 "__addsf3", 0x7ffff183ed2c "__adddf3", 0x7ffff183ed35 "__addxf3", 0x7ffff183ed3e "__addtf3", 0x7ffff183ed47 "__gcc_qadd", 0x7ffff183ed52 "__subsf3", 0x7ffff183ed5b "__subdf3", 0x7ffff183ed64 "__subxf3", 0x7ffff183ed6d "__subtf3", 0x7ffff183ed76 "__gcc_qsub", 0x7ffff183ed81 "__mulsf3", 0x7ffff183ed8a "__muldf3", 0x7ffff183ed93 "__mulxf3", 0x7ffff183ed9c "__multf3", 0x7ffff183eda5 "__gcc_qmul", 0x7ffff183edb0 "__divsf3", 0x7ffff183edb9 "__divdf3", 0x7ffff183edc2 "__divxf3", 0x7ffff183edcb "__divtf3", 0x7ffff183edd4 "__gcc_qdiv", 0x7ffff183eddf "fmodf", 0x7ffff183ede5 "fmod", 0x7ffff183edea "fmodl", 0x7ffff183edea "fmodl", 0x7ffff183edea "fmodl", 0x7ffff183edf0 "fmaf", 0x7ffff183edf5 "fma", 0x7ffff183edf9 "fmal", 0x7ffff183edf9 "fmal", 0x7ffff183edf9 "fmal", 0x7ffff183edfe "__powisf2", 0x7ffff183ee08 "__powidf2", 0x7ffff183ee12 "__powixf2", 0x7ffff183ee1c "__powitf2", 0x7ffff183ee1c "__powitf2", 0x7ffff183ee26 "sqrtf", 0x7ffff183ee2c "sqrt", 0x7ffff183ee31 "sqrtl", 0x7ffff183ee31 "sqrtl", 0x7ffff183ee31 "sqrtl", 0x7ffff183ee37 "logf", 0x7ffff183ee3c "log", 0x7ffff183ee40 "logl", 0x7ffff183ee40 "logl", 0x7ffff183ee40 "logl", 0x7ffff183ee45 "log2f", 0x7ffff183ee4b "log2", 0x7ffff183ee50 "log2l", 0x7ffff183ee50 "log2l", 0x7ffff183ee50 "log2l", 0x7ffff183ee56 "log10f", 0x7ffff183ee5d "log10", 0x7ffff183ee63 "log10l", 0x7ffff183ee63 "log10l", 0x7ffff183ee63 "log10l", 0x7ffff183ee6a "expf", 0x7ffff183ee6f "exp", 0x7ffff183ee73 "expl", 0x7ffff183ee73 "expl", 0x7ffff183ee73 "expl", 0x7ffff183ee78 "exp2f", 0x7ffff183ee7e "exp2", 0x7ffff183ee83 "exp2l", 0x7ffff183ee83 "exp2l", 0x7ffff183ee83 "exp2l", 0x7ffff183ee89 "sinf", 0x7ffff183ee8e "sin", 0x7ffff183ee92 "sinl", 0x7ffff183ee92 "sinl", 0x7ffff183ee92 "sinl", 0x7ffff183ee97 "cosf", 0x7ffff183ee9c "cos", 0x7ffff183eea0 "cosl", 0x7ffff183eea0 "cosl", 0x7ffff183eea0 "cosl", 0x0, 0x0, 0x0, 0x0, 0x0, 0x7ffff183eea5 "powf", 0x7ffff183eeaa "pow", 0x7ffff183eeae "powl", 0x7ffff183eeae "powl", 0x7ffff183eeae "powl", 0x7ffff183eeb3 "ceilf", 0x7ffff183eeb9 "ceil", 0x7ffff183eebe "ceill", 0x7ffff183eebe "ceill", 0x7ffff183eebe "ceill", 0x7ffff183eec4 "truncf", 0x7ffff183eecb "trunc", 0x7ffff183eed1 "truncl", 0x7ffff183eed1 "truncl", 0x7ffff183eed1 "truncl", 0x7ffff183eed8 "rintf", 0x7ffff183eede "rint", 0x7ffff183eee3 "rintl", 0x7ffff183eee3 "rintl", 0x7ffff183eee3 "rintl", 0x7ffff183eee9 "nearbyintf", 0x7ffff183eef4 "nearbyint", 0x7ffff183eefe "nearbyintl", 0x7ffff183eefe "nearbyintl", 0x7ffff183eefe "nearbyintl", 0x7ffff183ef09 "floorf", 0x7ffff183ef10 "floor", 0x7ffff183ef16 "floorl", 0x7ffff183ef16 "floorl", 0x7ffff183ef16 "floorl", 0x7ffff183ef1d "copysignf", 0x7ffff183ef27 "copysign", 0x7ffff183ef30 "copysignl", 0x7ffff183ef30 "copysignl", 0x7ffff183ef30 "copysignl", 0x7ffff183ef3a "__extenddftf2", 0x7ffff183ef48 "__extendsftf2", 0x7ffff183ef56 "__extendsfdf2", 0x7ffff183ef64 "__gnu_h2f_ieee", 0x7ffff183ef73 "__gnu_f2h_ieee", 0x7ffff183ef82 "__truncdfsf2", 0x7ffff183ef8f "__truncxfsf2", 0x7ffff183ef9c "__trunctfsf2", 0x7ffff183ef9c "__trunctfsf2", 0x7ffff183efa9 "__truncxfdf2", 0x7ffff183efb6 "__trunctfdf2", 0x7ffff183efb6 "__trunctfdf2", 0x7ffff183efc3 "__fixsfqi", 0x7ffff183efcd "__fixsfhi", 0x7ffff183efd7 "__fixsfsi", 0x7ffff183efe1 "__fixsfdi", 0x7ffff183efeb "__fixsfti", 0x7ffff183eff5 "__fixdfqi", 0x7ffff183efff "__fixdfhi", 0x7ffff183f009 "__fixdfsi", 0x7ffff183f013 "__fixdfdi", 0x7ffff183f01d "__fixdfti", 0x7ffff183f027 "__fixxfsi", 0x7ffff183f031 "__fixxfdi", 0x7ffff183f03b "__fixxfti", 0x7ffff183f045 "__fixtfsi", 0x7ffff183f04f "__fixtfdi", 0x7ffff183f059 "__fixtfti", 0x7ffff183f045 "__fixtfsi", 0x7ffff183f04f "__fixtfdi", 0x7ffff183f059 "__fixtfti", 0x7ffff183f063 "__fixunssfqi", 0x7ffff183f070 "__fixunssfhi"...}, CmpLibcallCCs = {404232216 , llvm::ISD::SETEQ, llvm::ISD::SETEQ, llvm::ISD::SETEQ, llvm::ISD::SETNE, llvm::ISD::SETNE, llvm::ISD::SETNE, llvm::ISD::SETGE, llvm::ISD::SETGE, llvm::ISD::SETGE, llvm::ISD::SETLT, llvm::ISD::SETLT, llvm::ISD::SETLT, llvm::ISD::SETLE, llvm::ISD::SETLE, llvm::ISD::SETLE, llvm::ISD::SETGT, llvm::ISD::SETGT, llvm::ISD::SETGT, llvm::ISD::SETNE, llvm::ISD::SETNE, llvm::ISD::SETNE, llvm::ISD::SETEQ, llvm::ISD::SETEQ, llvm::ISD::SETEQ, 404232216 }, LibcallCallingConvs = {llvm::CallingConv::C }, MaxStoresPerMemset = 4096, MaxStoresPerMemsetOptSize = 4, MaxStoresPerMemcpy = 4096, MaxStoresPerMemcpyOptSize = 4, MaxStoresPerMemmove = 4096, MaxStoresPerMemmoveOptSize = 4, PredictableSelectIsExpensive = false}, }, } #2 0x00007ffff32c8b8f in llvm::SelectionDAGISel::DoInstructionSelection (this=0x704760) at /work_local/llvm/lib/CodeGen/SelectionDAG/SelectionDAGISel.cpp:810 No locals. #3 0x00007ffff32c8603 in llvm::SelectionDAGISel::CodeGenAndEmitDAG (this=0x704760) at /work_local/llvm/lib/CodeGen/SelectionDAG/SelectionDAGISel.cpp:684 T = { = {T = 0x0}, } Scheduler = 0x7fffffffcf50 LastMBB = 0x7fffffffcf50 GroupName = {static npos = , _M_dataplus = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_p = 0x60a858 ""}} BlockName = {static npos = , _M_dataplus = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_p = 0x60a858 ""}} BlockNumber = -1 Changed = false FirstMBB = 0x7ffff3274ffe #4 0x00007ffff32c7c44 in llvm::SelectionDAGISel::SelectBasicBlock (this=0x704760, Begin=..., End=..., HadTailCall=@0x7fffffffd01f: false) at /work_local/llvm/lib/CodeGen/SelectionDAG/SelectionDAGISel.cpp:525 No locals. #5 0x00007ffff32c9982 in llvm::SelectionDAGISel::SelectAllBasicBlocks (this=0x704760, Fn=...) at /work_local/llvm/lib/CodeGen/SelectionDAG/SelectionDAGISel.cpp:1126 HadTailCall = false LLVMBB = 0xe50380 Begin = {> = {}, NodePtr = 0x6d0950} End = {> = {}, NodePtr = 0xe50380} BI = {> = {}, NodePtr = 0xe50380} I = {> = {}, current = {_M_current = 0xe17dc8}} E = {> = {}, current = {_M_current = 0xe17d90}} FastIS = 0x0 RPOT = {Blocks = { >> = {_M_impl = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_start = 0xe17d90, _M_finish = 0xe17dd8, _M_end_of_storage = 0xe17e10}}, }} #6 0x00007ffff32c6f50 in llvm::SelectionDAGISel::runOnMachineFunction (this=0x704760, mf=...) at /work_local/llvm/lib/CodeGen/SelectionDAG/SelectionDAGISel.cpp:386 Fn = @0xe59610: { = { = { = { = {_vptr.Value = 0x7fffef82e050, SubclassID = 2 '\002', HasValueHandle = 1 '\001', SubclassOptionalData = 0 '\000', SubclassData = 0, VTy = 0xe4d750, UseList = 0x0, Name = 0xe558c0, static MaximumAlignment = 536870912}, OperandList = 0x0, NumOperands = 0}, }, Linkage = llvm::GlobalValue::ExternalLinkage, Visibility = 0, Alignment = 0, UnnamedAddr = 0, Parent = 0x6a1630, Section = {static npos = , _M_dataplus = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_p = 0x60a858 ""}}}, > = {> = {Prev = 0x0}, Next = 0xe53350}, BasicBlocks = {> = {> = {> = {> = {}, > = {}, > = {}, }, }, Sentinel = {Prev = 0xe4ef10}}, Head = 0xe4bde0}, ArgumentList = {> = {> = {> = {> = {}, > = {}, > = {}, }, }, Sentinel = {Prev = 0xe501a0}}, Head = 0xe50b90}, SymTab = 0xe52ec0, AttributeSets = {pImpl = 0xe4bd30}} TRI = @0xe52b50: { = {Desc = 0x7ffff58d40e0, NumRegs = 3819, RAReg = 0, PCReg = 0, Classes = 0x7ffff5b07760, NumClasses = 49, NumRegUnits = 2251, RegUnitRoots = 0x7ffff58ea700, DiffLists = 0x7ffff58be500, RegStrings = 0x7ffff58c73c0 "SGPR100", SubRegIndices = 0x7ffff58c7240, NumSubRegIndices = 68, RegEncodingTable = 0x7ffff58ed1c0, L2DwarfRegsSize = 42, EHL2DwarfRegsSize = 42, Dwarf2LRegsSize = 42, EHDwarf2LRegsSize = 42, L2DwarfRegs = 0x7ffff58ecf00, EHL2DwarfRegs = 0x7ffff58ed060, Dwarf2LRegs = 0x7ffff58ecc40, EHDwarf2LRegs = 0x7ffff58ecda0, L2SEHRegs = { >, unsigned int, int, llvm::DenseMapInfo >> = {}, Buckets = 0x0, NumEntries = 0, NumTombstones = 0, NumBuckets = 0}}, _vptr.TargetRegisterInfo = 0x7ffff5e59930, InfoDesc = 0x7ffff5c134a0, SubRegIndexNames = 0x7ffff5e55580, SubRegIndexLaneMasks = 0x7ffff5c128a0, RegClassBegin = 0x7ffff5e581e0, RegClassEnd = 0x7ffff5e58368, CoveringLanes = 4294901760} LiveInMap = { >, unsigned int, unsigned int, llvm::DenseMapInfo >> = {}, Buckets = 0x6a1630, NumEntries = 14771424, NumTombstones = 0, NumBuckets = 4294955856} TII = @0xe52a30: { = {Desc = 0x7ffff5afb140, InstrNameIndices = 0x7ffff58bce00, InstrNameData = 0x7ffff58b8fa0 "CF_TC_R600", NumOpcodes = 1025}, _vptr.TargetInstrInfo = 0x7ffff5e592b0, CallFrameSetupOpcode = 0, CallFrameDestroyOpcode = 0} ST = @0x99c348: { = {TargetTriple = {static npos = , _M_dataplus = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_p = 0x6d2a28 "r600--"}}, ProcFeatures = 0x7ffff5b07180, ProcDesc = 0x7ffff5b07340, ProcSchedModels = 0x7ffff5b07600, WriteProcResTable = 0x7ffff58be4e0, WriteLatencyTable = 0x7ffff58be4e8, ReadAdvanceTable = 0x7ffff58be4f0, CPUSchedModel = 0x7ffff5b08740, Stages = 0x7ffff58be440, OperandCycles = 0x7ffff58be4d0, ForwardingPaths = 0x7ffff58be4d8, NumFeatures = 14, NumProcs = 22, FeatureBits = 66}, _vptr.TargetSubtargetInfo = 0x7ffff5e565f0} EntryMBB = 0xe504a0 MFI = 0x4 MRI = @0x7fffffffda50: {TRI = 0x0, IsSSA = false, TracksLiveness = false, VRegInfo = {storage_ = {, std::allocator > >> = { _M_impl = { >> = {<__gnu_cxx::new_allocator >> = {}, }, _M_start = 0x0, _M_finish = 0x0, _M_end_of_storage = 0x7fffffffda90}}, }, nullVal_ = {first = 0x7fffef47bf24, second = 0x6df9}, toIndex_ = {> = {}, }}, RegAllocHints = { storage_ = {, std::allocator > >> = { _M_impl = { >> = {<__gnu_cxx::new_allocator >> = {}, }, _M_start = 0xe5a9b0, _M_finish = 0x7fffeeb5cfb9, _M_end_of_storage = 0x660}}, }, nullVal_ = { first = 8520148, second = 0}, toIndex_ = {> = {}, }}, PhysRegUseDefLists = 0x1, UsedRegUnits = {Bits = 0x63b130, Size = 8520144, Capacity = 0}, UsedPhysRegMask = {Bits = 0x8201d4, Size = 4004860057, Capacity = 32767}, ReservedRegs = {Bits = 0x0, Size = 9039744, Capacity = 0}, LiveIns = {, std::allocator > >> = { _M_impl = { >> = {<__gnu_cxx::new_allocator >> = {}, }, _M_start = 0x820870, _M_finish = 0x89ef80, _M_end_of_storage = 0x0}}, }} #7 0x00007ffff170b4e9 in llvm::MachineFunctionPass::runOnFunction (this=0x704760, F=...) at /work_local/llvm/lib/CodeGen/MachineFunctionPass.cpp:33 MF = @0xe13d20: {Fn = 0xe59610, Target = @0x99c2e0, Ctx = @0x703ab0, MMI = @0x703a90, GMI = 0x6fe1c0, RegInfo = 0xe180f0, MFInfo = 0xe18ab0, FrameInfo = 0xe181b0, ConstantPool = 0xe18440, JumpTableInfo = 0x0, MBBNumbering = { >> = {_M_impl = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_start = 0xe13fa0, _M_finish = 0xe13fe8, _M_end_of_storage = 0xe14020}}, }, Allocator = {SlabSize = 4096, SizeThreshold = 4096, Allocator = @0x7fffef355100, CurSlab = 0xe45500, CurPtr = 0xe464d0 "", End = 0xe46500 "X\250`", BytesAllocated = 8088, static DefaultSlabAllocator = { = { _vptr.SlabAllocator = 0x7fffef34c590}, Allocator = {}}}, InstructionRecycler = { FreeList = {> = {> = {> = {}, > = {}, > = {}, }, Sentinel = {Prev = 0x1, Next = 0xe14412}}, Head = 0xe13dc0}}, OperandRecycler = { Bucket = {::FreeList*>> = {::FreeList*, true>> = {::FreeList*, void>> = { = {BeginX = 0xe13df0, EndX = 0xe13df0, CapacityX = 0xe13e30}, FirstEl = {> = {buffer = "\000\000\000\000\000\000\000"}, }}, }, }, Storage = {InlineElts = {{> = { buffer = "\000\000\000\000\000\000\000"}, }, {> = {buffer = "\000\000\000\000\000\000\000"}, }, {> = {buffer = "\a\000\000\000\b\000\000"}, }, {> = {buffer = "\240\ae\000\000\000\000"}, }, {> = {buffer = "A\000\000\000\000\000\000"}, }, {> = {buffer = "\020C\341\000\000\000\000"}, }, {> = {buffer = "\000\033\341\000\000\000\000"}, }}}}}, BasicBlockRecycler = { FreeList = {> = {> = {> = {}, > = {}, > = {}, }, Sentinel = {Prev = 0x1, Next = 0xe4be42}}, Head = 0xe13e30}}, BasicBlocks = { >> = {> = {> = {> = {}, > = {}, > = {}, }, Sentinel = {Prev = 0xe18a00}}, Head = 0xe18480}, }, FunctionNumber = 0, Alignment = 0, ExposesReturnsTwice = 64, HasMSInlineAsm = false} #8 0x00007fffef54ef02 in llvm::FPPassManager::runOnFunction (this=0x6de7c0, F=...) at /work_local/llvm/lib/IR/PassManager.cpp:1530 X = { = {_vptr.PrettyStackTraceEntry = 0x7fffef82ff30, NextEntry = 0x7fffffffd650}, P = 0x704760, V = 0xe59610, M = 0x0} PassTimer = {T = 0x0} FP = 0x704760 LocalChanged = false Index = 20 Changed = true #9 0x00007fffef54f128 in llvm::FPPassManager::runOnModule (this=0x6de7c0, M=...) at /work_local/llvm/lib/IR/PassManager.cpp:1550 I = {> = {}, NodePtr = 0xe59610} E = {> = {}, NodePtr = 0x6a1608} Changed = false #10 0x00007fffef54f485 in llvm::MPPassManager::runOnModule (this=0xe57b10, M=...) at /work_local/llvm/lib/IR/PassManager.cpp:1608 X = { = {_vptr.PrettyStackTraceEntry = 0x7fffef82ff30, NextEntry = 0x0}, P = 0x6de7c0, V = 0x0, M = 0x6a1630} PassTimer = {T = 0x0} MP = 0x6de7c0 LocalChanged = false Index = 0 Changed = true #11 0x00007fffef54f9d3 in llvm::PassManagerImpl::run (this=0x6fe290, M=...) at /work_local/llvm/lib/IR/PassManager.cpp:1703 Index = 0 Changed = false IPV = @0x6fe508: {> = {> = { = {BeginX = 0x704340, EndX = 0x704390, CapacityX = 0x7043c8}, FirstEl = {> = { buffer = "\340'\345\000\000\000\000"}, }}, }, } #12 0x00007fffef54fbe5 in llvm::PassManager::run (this=0x7fffffffd810, M=...) at /work_local/llvm/lib/IR/PassManager.cpp:1738 No locals. #13 0x00007fffefdf4a9c in LLVMTargetMachineEmit (T=0x99c2e0, M=0x6a1630, OS=..., codegen=LLVMObjectFile, ErrorMessage=0x7fffffffd978) at /work_local/llvm/lib/Target/TargetMachineC.cpp:194 td = 0x99c430 ft = llvm::TargetMachine::CGFT_ObjectFile TM = 0x99c2e0 Mod = 0x6a1630 pass = { = {_vptr.PassManagerBase = 0x7fffef82f8d0}, PM = 0x6fe290} error = {static npos = , _M_dataplus = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_p = 0x60a858 ""}} #14 0x00007fffefdf4c8d in LLVMTargetMachineEmitToMemoryBuffer (T=0x99c2e0, M=0x6a1630, codegen=LLVMObjectFile, ErrorMessage=0x7fffffffd978, OutMemBuf=0x7fffffffd980) at /work_local/llvm/lib/Target/TargetMachineC.cpp:220 CodeString = {static npos = , _M_dataplus = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_p = 0x60a858 ""}} OStream = { = {_vptr.raw_ostream = 0x7fffef34e290, OutBufStart = 0x0, OutBufEnd = 0x0, OutBufCur = 0x0, BufferMode = llvm::raw_ostream::Unbuffered}, OS = @0x7fffffffd910} Data = @0xffffd940: Out = { = {_vptr.raw_ostream = 0x7fffef34d6d0, OutBufStart = 0x6fc1b0 "\030\016;\020;", OutBufEnd = 0x6fe1b0 "\310\341o", OutBufCur = 0x6fc1b0 "\030\016;\020;", BufferMode = llvm::raw_ostream::InternalBuffer}, static DELETE_STREAM = true, static PRESERVE_STREAM = false, TheStream = 0x7fffffffd890, DeleteStream = false, Position = {first = 0, second = 0}, Scanned = 0x0} Result = false #15 0x00007fffeeb5cc5e in radeon_llvm_compile (M=M@entry=0x6a1630, binary=binary@entry=0x7fffffffda50, gpu_family=0x99c2e0 "\260h\345\365\377\177", dump=dump@entry=0) at ../../../../../src/gallium/drivers/radeon/radeon_llvm_emit.c:124 target = 0x7ffff58a1340 tm = 0x99c2e0 cpu = "pitcairn", '\000' fs = '\000' err = out_buffer = buffer_size = buffer_data = triple = "r600--" elf_buffer = elf = section = 0x0 section_str_index = r = #16 0x00007fffeeb527cd in si_compile_llvm (rctx=rctx@entry=0x615840, shader=0xe573c0, mod=0x6a1630) at ../../../../../src/gallium/drivers/radeonsi/radeonsi_shader.c:1168 i = ptr = dump = false binary = {code = 0x0, code_size = 0, config = 0x0, config_size = 0} #17 0x00007fffeeb55099 in radeonsi_create_compute_state (ctx=0x615840, cso=) at ../../../../../src/gallium/drivers/radeonsi/radeonsi_compute.c:50 mod = rctx = 0x615840 program = 0x63b130 header = 0x8201d0 code = 0x8201d4 "BC\300\336!\f" i = #18 0x00007ffff6993197 in _cl_kernel::exec_context::bind (this=this@entry=0x820870, __q=0x0, __q@entry=0x6157f0) at ../../../../../src/gallium/state_trackers/clover/core/kernel.cpp:159 No locals. #19 0x00007ffff6993e6e in _cl_kernel::launch (this=0x820840, q=..., grid_offset=..., grid_size=..., block_size=...) at ../../../../../src/gallium/state_trackers/clover/core/kernel.cpp:68 st = g_handles = { >> = {_M_impl = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_start = 0x20, _M_finish = 0x7fffffffdc60, _M_end_of_storage = 0x11f0}}, } #20 0x00007ffff698f90c in trigger (this=0xe52550) at ../../../../../src/gallium/state_trackers/clover/core/event.cpp:43 No locals. #21 _cl_event::trigger (this=0xe52550) at ../../../../../src/gallium/state_trackers/clover/core/event.cpp:41 No locals. #22 0x00007ffff6990117 in clover::hard_event::hard_event(_cl_command_queue&, unsigned int, std::vector<_cl_event*, std::allocator<_cl_event*> >, std::function) (this=0xe52550, q=..., command=4592, deps=..., action=...) at ../../../../../src/gallium/state_trackers/clover/core/event.cpp:82 No locals. #23 0x00007ffff69a8567 in clEnqueueNDRangeKernel (q=0x6157f0, kern=0x820840, dims=, pgrid_offset=, pgrid_size=, pblock_size=, num_deps=0, deps=0x0, ev=0x0) at ../../../../../src/gallium/state_trackers/clover/api/kernel.cpp:286 grid_offset = { >> = {_M_impl = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_start = 0xe03cf0, _M_finish = 0xe03d00, _M_end_of_storage = 0xe03d00}}, } grid_size = { >> = {_M_impl = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_start = 0x62cc90, _M_finish = 0x62cca0, _M_end_of_storage = 0x62cca0}}, } block_size = { >> = {_M_impl = {> = {<__gnu_cxx::new_allocator> = {}, }, _M_start = 0xe04f90, _M_finish = 0xe04fa0, _M_end_of_storage = 0xe04fa0}}, } hev = #24 0x00000000004066d5 in Laplace::run(cl::CommandQueue) () No symbol table info available. #25 0x000000000040275a in main () No symbol table info available. A debugging session is active. Inferior 1 [process 28153] will be killed. Quit anyway? (y or n)