Searched refs:GetBasePointer (Results 1 – 10 of 10) sorted by relevance
| /external/tensorflow/tensorflow/compiler/xla/service/llvm_ir/ |
| D | tuple_ops.cc | 43 b->CreateLoad(pred.GetBasePointer(), "load_predicate_value"); in EmitTupleSelect()
54 llvm::Value* dst = select.GetBasePointer(); in EmitTupleSelect()
68 cast, b->CreateInBoundsGEP(tuple.GetBasePointer(), in EmitTuple()
80 [](const llvm_ir::IrArray& buffer) { return buffer.GetBasePointer(); }); in EmitTuple()
|
| D | ir_array.h | 223 llvm::Value* GetBasePointer() const { return base_ptr_; } in GetBasePointer() function
|
| /external/tensorflow/tensorflow/compiler/xla/service/cpu/ |
| D | dot_op_emitter.cc | 256 llvm::Value* operand_ptrs[] = {lhs_array_.GetBasePointer(), in EmitLinalgMatmul()
257 rhs_array_.GetBasePointer()}; in EmitLinalgMatmul()
258 llvm::Value* target_ptr = target_array_.GetBasePointer(); in EmitLinalgMatmul()
346 llvm::Value* lhs = lhs_array_.GetBasePointer(); in EmitTiledLlvmIrGemm()
347 llvm::Value* rhs = rhs_array_.GetBasePointer(); in EmitTiledLlvmIrGemm()
348 llvm::Value* target = target_array_.GetBasePointer(); in EmitTiledLlvmIrGemm()
453 llvm::Value* result_op = target_array_.GetBasePointer(); in EmitTiledLlvmIrGemv()
455 swap_operands ? rhs_array_.GetBasePointer() : lhs_array_.GetBasePointer(); in EmitTiledLlvmIrGemv()
457 swap_operands ? lhs_array_.GetBasePointer() : rhs_array_.GetBasePointer(); in EmitTiledLlvmIrGemv()
479 /*addend=*/addend_array_ ? addend_array_->GetBasePointer() : nullptr, in EmitTiledLlvmIrGemv()
[all …]
|
| D | ir_emitter.cc | 125 Load(root_value.GetBasePointer(), "load_ret_value"); in EmitThreadLocalFunctionEpilogue()
127 BitCast(out_parameter, root_value.GetBasePointer()->getType())); in EmitThreadLocalFunctionEpilogue()
147 root_value.GetBasePointer(), &b_); in EmitThreadLocalFunctionEpilogue()
857 << llvm_ir::DumpToString(*lhs_array.GetBasePointer()); in HandleDot()
859 << llvm_ir::DumpToString(*rhs_array.GetBasePointer()); in HandleDot()
861 << llvm_ir::DumpToString(*target_array.GetBasePointer()); in HandleDot()
2206 tuple_operand_ptrs.push_back(data_array.GetBasePointer()); in HandlePadToStatic()
2661 GetIrArrayFor(branch_index).GetBasePointer(), "load_predicate_value"); in HandleConditional()
2696 GetIrArrayFor(branch_index).GetBasePointer(), "load_branch_index_value"); in HandleConditional()
3159 tuple_operand_ptrs.push_back(output_arrays[i].GetBasePointer()); in EmitTargetElementLoop()
|
| /external/tensorflow/tensorflow/compiler/xla/service/gpu/ |
| D | ir_emitter.h | 131 llvm::Value* GetBasePointer(const HloInstruction& inst,
133 return bindings_.GetBasePointer(inst, shape_index);
|
| D | ir_emitter.cc | 177 bindings_.BindHloToIrValue(*add_dependency, GetBasePointer(*operand)); in HandleAddDependency()
191 /*alignment=*/1, GetBasePointer(*operand), &b_)); in HandleGetTupleElement()
218 base_ptrs.push_back(GetBasePointer(*operand)); in HandleTuple()
592 operand_addresses.push_back(GetBasePointer(*operand)); in HandleCall()
595 GetBasePointer(*call)); in HandleCall()
|
| D | hlo_to_ir_bindings.h | 71 llvm::Value* GetBasePointer(const HloInstruction& hlo,
|
| D | ir_emitter_nested.cc | 127 llvm::Value* root_value = bindings_.GetBasePointer(*root_instruction); in CodegenNestedComputation()
|
| D | hlo_to_ir_bindings.cc | 194 llvm::Value* base_ptr = GetBasePointer(hlo, shape_index); in GetIrArray()
|
| D | ir_emitter_unnested.cc | 838 llvm::Value* source_buffer = source_array.GetBasePointer(); in EmitPadToStaticFromMlir()
878 output_dim_arrays[dim_index].GetBasePointer(); in EmitPadToStaticFromMlir()
974 llvm::Value* dest_buffer = data_array.GetBasePointer(); in EmitSliceToDynamicFromMlir()
982 llvm::Value* source_buffer = ir_arrays[i].GetBasePointer(); in EmitSliceToDynamicFromMlir()
|