Home
last modified time | relevance | path

Searched refs:instr1 (Results 1 – 25 of 28) sorted by relevance

12

/external/tensorflow/tensorflow/compiler/xla/service/
Dmulti_output_fusion.cc132 HloInstruction* MultiOutputFusion::Fuse(HloInstruction* instr1, in Fuse() argument
134 HloInstruction* remaining = instr1; in Fuse()
193 MultiOutputFusion::GetNewFusibles(HloInstruction* instr1, in GetNewFusibles() argument
195 HloInstruction* fusion = instr1; in GetNewFusibles()
197 if (is_fused(instr1)) { in GetNewFusibles()
199 fused = instr1; in GetNewFusibles()
240 void MultiOutputFusion::UpdateBeforeFuse(HloInstruction* instr1, in UpdateBeforeFuse() argument
242 HloInstruction* fusion = instr1; in UpdateBeforeFuse()
244 if (is_fused(instr1)) { in UpdateBeforeFuse()
246 fused = instr1; in UpdateBeforeFuse()
[all …]
Dmulti_output_fusion.h65 virtual bool ShapesCompatibleForFusion(HloInstruction* instr1,
74 virtual int64 GetProfit(HloInstruction* instr1, HloInstruction* instr2) = 0;
80 virtual bool LegalToFuse(HloInstruction* instr1, HloInstruction* instr2);
84 bool LegalToFuseMainConstraints(HloInstruction* instr1,
89 virtual HloInstruction* Fuse(HloInstruction* instr1, HloInstruction* instr2);
102 HloInstruction* instr1, HloInstruction* instr2,
118 HloInstruction* instr1, HloInstruction* instr2);
136 HloInstruction* instr1; member
139 ToBeFused(HloInstruction* instr1, HloInstruction* instr2, int64 score) in ToBeFused()
140 : instr1(instr1), instr2(instr2), score(score) {} in ToBeFused()
[all …]
Dhlo_instruction_test.cc1771 auto instr1 = HloInstruction::CreateCustomCall(ShapeUtil::MakeShape(F32, {}), in TEST_F() local
1774 auto instr2 = instr1->Clone(); in TEST_F()
1775 EXPECT_TRUE(instr1->Identical(*instr2)); in TEST_F()
1778 instr1->set_window(w); in TEST_F()
1779 EXPECT_FALSE(instr1->Identical(*instr2)); in TEST_F()
1783 auto instr1 = HloInstruction::CreateCustomCall(ShapeUtil::MakeShape(F32, {}), in TEST_F() local
1786 auto instr2 = instr1->Clone(); in TEST_F()
1787 EXPECT_TRUE(instr1->Identical(*instr2)); in TEST_F()
1791 instr1->set_convolution_dimension_numbers(dnums); in TEST_F()
1792 EXPECT_FALSE(instr1->Identical(*instr2)); in TEST_F()
[all …]
Dhlo_verifier.cc1272 Status CheckSameChannel(const HloInstruction* instr1, in CheckSameChannel() argument
1274 if (instr1->channel_id() != instr2->channel_id()) { in CheckSameChannel()
1278 instr1->ToString(), *instr1->channel_id(), instr2->ToString(), in CheckSameChannel()
1287 Status CheckSameIsHostTransfer(const HloInstruction* instr1, in CheckSameIsHostTransfer() argument
1290 DynCast<const HloSendRecvInstruction>(instr1); in CheckSameIsHostTransfer()
1300 instr1->ToString(), instr2->ToString()); in CheckSameIsHostTransfer()
/external/compiler-rt/test/profile/
Dinstrprof-shared.test5 2. libt-no-instr1.so is not instrumented
18 RUN: %clang -o %t.d/libt-no-instr1.so -fPIC -shared %S/Inputs/instrprof-shared-lib.c
23 RUN: %clang_profgen -o %t-instr-no-instr1 -L%t.d -rpath %t.d -lt-no-instr1 %S/Inputs/instrprof-sha…
25 RUN: %clang -o %t-no-instr1-instr -L%t.d -rpath %t.d -lt-instr %S/Inputs/instrprof-shared-main.c
26 RUN: %clang -o %t-no-instr1-no-instr1 -L%t.d -rpath %t.d -lt-no-instr1 %S/Inputs/instrprof-shared-…
27 RUN: %clang -o %t-no-instr1-no-instr2 -L%t.d -rpath %t.d -lt-no-instr2 %S/Inputs/instrprof-shared-…
30 RUN: %clang -o %t-no-instr2-no-instr1 -L%t.d -rpath %t.d -lt-no-instr1 %t.d/instrprof-shared-main-…
34 RUN: env LLVM_PROFILE_FILE=%t-instr-no-instr1.profraw %run %t-instr-no-instr1
36 RUN: env LLVM_PROFILE_FILE=%t-no-instr1-instr.profraw %run %t-no-instr1-instr
38 RUN: env LLVM_PROFILE_FILE=%t-no-instr1-no-instr1.profraw %run %t-no-instr1-no-instr1
[all …]
/external/tensorflow/tensorflow/compiler/xla/service/gpu/
Dmulti_output_fusion.cc40 bool GpuMultiOutputFusion::ShapesCompatibleForFusion(HloInstruction* instr1, in ShapesCompatibleForFusion() argument
42 return ShapesCompatibleForMultiOutputFusion(*instr1, *instr2); in ShapesCompatibleForFusion()
49 int64 GpuMultiOutputFusion::GetProfit(HloInstruction* instr1, in GetProfit() argument
52 for (auto instr : instr1->operands()) { in GetProfit()
63 VLOG(2) << "Fusing instr1=" << instr1->name() << " instr2=" << instr2->name() in GetProfit()
68 bool GpuMultiOutputFusion::LegalToFuse(HloInstruction* instr1, in LegalToFuse() argument
70 if (!MultiOutputFusion::LegalToFuse(instr1, instr2)) { in LegalToFuse()
79 CHECK(instr1->opcode() == HloOpcode::kFusion); in LegalToFuse()
81 instr1->fusion_kind() != instr2->fusion_kind()) || in LegalToFuse()
83 instr1->IsLoopFusion())) { in LegalToFuse()
[all …]
Dmulti_output_fusion.h33 bool ShapesCompatibleForFusion(HloInstruction* instr1,
44 int64 GetProfit(HloInstruction* instr1, HloInstruction* instr2) override;
47 bool LegalToFuse(HloInstruction* instr1, HloInstruction* instr2) override;
Dgpu_fusible.cc117 bool ShapesCompatibleForMultiOutputFusion(const HloInstruction& instr1, in ShapesCompatibleForMultiOutputFusion() argument
155 auto* instr_1 = get_real_hero(&instr1); in ShapesCompatibleForMultiOutputFusion()
298 bool FusionWouldBeTooLarge(const HloInstruction& instr1, in FusionWouldBeTooLarge() argument
315 int64 num_output_buffers = ShapeUtil::SubshapeCount(instr1.shape()) + in FusionWouldBeTooLarge()
325 if (instr1.operand_count() + instr2.operand_count() - 1 + in FusionWouldBeTooLarge()
332 absl::flat_hash_set<const HloInstruction*> operands(instr1.operands().begin(), in FusionWouldBeTooLarge()
333 instr1.operands().end()); in FusionWouldBeTooLarge()
337 operands.erase(&instr1); in FusionWouldBeTooLarge()
Dgpu_fusible.h67 bool FusionWouldBeTooLarge(const HloInstruction& instr1,
82 bool ShapesCompatibleForMultiOutputFusion(const HloInstruction& instr1,
/external/v8/src/codegen/mips/
Dassembler-mips-inl.h125 Instr instr1 = Assembler::instr_at(pc + 0 * kInstrSize); in set_target_internal_reference_encoded_at() local
127 DCHECK(Assembler::IsLui(instr1)); in set_target_internal_reference_encoded_at()
129 instr1 &= ~kImm16Mask; in set_target_internal_reference_encoded_at()
138 Assembler::instr_at_put(pc + 0 * kInstrSize, instr1 | lui_offset_u); in set_target_internal_reference_encoded_at()
142 PatchLuiOriImmediate(pc, imm, instr1, 0 * kInstrSize, instr2, in set_target_internal_reference_encoded_at()
211 Instr instr1 = Assembler::instr_at(pc_ + 0 * kInstrSize); in target_internal_reference() local
213 DCHECK(Assembler::IsLui(instr1)); in target_internal_reference()
217 Assembler::CreateTargetAddress(instr1, instr2)); in target_internal_reference()
219 return static_cast<Address>(Assembler::GetLuiOriImmediate(instr1, instr2)); in target_internal_reference()
262 Instr instr1 = instr_at(pc); in relative_code_target_object_handle_at() local
[all …]
Dassembler-mips.cc854 Instr instr1 = instr_at(pos + 0 * kInstrSize); in target_at() local
859 imm = CreateTargetAddress(instr1, instr2); in target_at()
861 imm = GetLuiOriImmediate(instr1, instr2); in target_at()
978 Instr instr1 = instr_at(pos + 0 * kInstrSize); in target_at_put() local
983 DCHECK(IsLui(instr1) && (IsJicOrJialc(instr2) || IsOri(instr2))); in target_at_put()
984 instr1 &= ~kImm16Mask; in target_at_put()
990 instr_at_put(pos + 0 * kInstrSize, instr1 | lui_offset_u); in target_at_put()
993 PatchLuiOriImmediate(pos, imm, instr1, 0 * kInstrSize, instr2, in target_at_put()
3465 Instr instr1 = instr_at(pc + 0 * kInstrSize); in MSA_BIT_LIST() local
3470 imm = CreateTargetAddress(instr1, instr2); in MSA_BIT_LIST()
[all …]
Dassembler-mips.h183 Instr instr1 = in pc_offset_for_safepoint() local
187 if (GetOpcodeField(instr1) != SPECIAL) { // instr1 == jialc. in pc_offset_for_safepoint()
188 DCHECK(IsMipsArchVariant(kMips32r6) && GetOpcodeField(instr1) == POP76 && in pc_offset_for_safepoint()
189 GetRs(instr1) == 0); in pc_offset_for_safepoint()
191 if (GetFunctionField(instr1) == SLL) { // instr1 == nop, instr2 == jalr. in pc_offset_for_safepoint()
195 DCHECK(GetFunctionField(instr1) == JALR); in pc_offset_for_safepoint()
1565 static int32_t GetLuiOriImmediate(Instr instr1, Instr instr2);
1811 static void PatchLuiOriImmediate(Address pc, int32_t imm, Instr instr1,
1814 void PatchLuiOriImmediate(int pc, int32_t imm, Instr instr1,
/external/v8/src/codegen/ppc/
Dassembler-ppc-inl.h283 Instr instr1 = instr_at(pc); in target_address_at() local
286 if (IsLis(instr1) && IsOri(instr2)) { in target_address_at()
291 uint64_t hi = (static_cast<uint32_t>((instr1 & kImm16Mask) << 16) | in target_address_at()
298 return static_cast<Address>(((instr1 & kImm16Mask) << 16) | in target_address_at()
397 Instr instr1 = instr_at(pc); in PatchConstantPoolAccessInstruction() local
399 instr1 &= ~kImm16Mask; in PatchConstantPoolAccessInstruction()
400 instr1 |= (hi_word & kImm16Mask); in PatchConstantPoolAccessInstruction()
403 instr_at_put(pc, instr1); in PatchConstantPoolAccessInstruction()
460 Instr instr1 = instr_at(pc); in set_target_address_at() local
463 if (IsLis(instr1) && IsOri(instr2)) { in set_target_address_at()
[all …]
Dassembler-ppc.cc336 bool Assembler::Is64BitLoadIntoR12(Instr instr1, Instr instr2, Instr instr3, in Is64BitLoadIntoR12() argument
344 return (((instr1 >> 16) == 0x3D80) && ((instr2 >> 16) == 0x618C) && in Is64BitLoadIntoR12()
350 bool Assembler::Is32BitLoadIntoR12(Instr instr1, Instr instr2) { in Is32BitLoadIntoR12() argument
354 return (((instr1 >> 16) == 0x3D80) && ((instr2 >> 16) == 0x618C)); in Is32BitLoadIntoR12()
Dassembler-ppc.h1143 static bool Is64BitLoadIntoR12(Instr instr1, Instr instr2, Instr instr3,
1146 static bool Is32BitLoadIntoR12(Instr instr1, Instr instr2);
/external/mesa3d/src/compiler/nir/
Dnir_opt_vectorize.c99 instrs_equal(const nir_instr *instr1, const nir_instr *instr2) in instrs_equal() argument
101 switch (instr1->type) { in instrs_equal()
103 nir_alu_instr *alu1 = nir_instr_as_alu(instr1); in instrs_equal()
165 instr_try_combine(struct nir_shader *nir, nir_instr *instr1, nir_instr *instr2, in instr_try_combine() argument
168 assert(instr1->type == nir_instr_type_alu); in instr_try_combine()
170 nir_alu_instr *alu1 = nir_instr_as_alu(instr1); in instr_try_combine()
189 nir_builder_init(&b, nir_cf_node_get_function(&instr1->block->cf_node)); in instr_try_combine()
190 b.cursor = nir_after_instr(instr1); in instr_try_combine()
312 nir_instr_remove(instr1); in instr_try_combine()
372 const nir_instr *instr1 = *(nir_instr **)util_dynarray_begin(arr1); in cmp_func() local
[all …]
Dnir_instr_set.c549 nir_instrs_equal(const nir_instr *instr1, const nir_instr *instr2) in nir_instrs_equal() argument
551 assert(instr_can_rewrite(instr1) && instr_can_rewrite(instr2)); in nir_instrs_equal()
553 if (instr1->type != instr2->type) in nir_instrs_equal()
556 switch (instr1->type) { in nir_instrs_equal()
558 nir_alu_instr *alu1 = nir_instr_as_alu(instr1); in nir_instrs_equal()
601 nir_deref_instr *deref1 = nir_instr_as_deref(instr1); in nir_instrs_equal()
645 nir_tex_instr *tex1 = nir_instr_as_tex(instr1); in nir_instrs_equal()
678 nir_load_const_instr *load1 = nir_instr_as_load_const(instr1); in nir_instrs_equal()
700 nir_phi_instr *phi1 = nir_instr_as_phi(instr1); in nir_instrs_equal()
720 nir_intrinsic_instr *intrinsic1 = nir_instr_as_intrinsic(instr1); in nir_instrs_equal()
/external/mesa3d/src/gallium/drivers/r600/sfn/
Dsfn_nir_vectorize_vs_inputs.c97 r600_io_access_same_var(const nir_instr *instr1, const nir_instr *instr2) in r600_io_access_same_var() argument
99 assert(instr1->type == nir_instr_type_intrinsic && in r600_io_access_same_var()
102 nir_intrinsic_instr *intr1 = nir_instr_as_intrinsic(instr1); in r600_io_access_same_var()
220 const nir_instr *instr1 = *(nir_instr **)util_dynarray_begin(arr1); in r600_cmp_func() local
223 return r600_io_access_same_var(instr1, instr2); in r600_cmp_func()
/external/v8/src/codegen/s390/
Dassembler-s390.cc433 bool Assembler::Is64BitLoadIntoIP(SixByteInstr instr1, SixByteInstr instr2) { in Is64BitLoadIntoIP() argument
435 return (((instr1 >> 32) == 0xC0C8) && ((instr2 >> 32) == 0xC0C9)); in Is64BitLoadIntoIP()
/external/tensorflow/tensorflow/compiler/tf2xla/
Dxla_compiler_test.cc761 auto instr1 = c1.instructions(j); in TEST_F() local
763 instr1.clear_name(); in TEST_F()
764 instr1.clear_id(); in TEST_F()
765 instr1.clear_operand_ids(); in TEST_F()
773 LOG(INFO) << "instr1 = " << instr1.DebugString(); in TEST_F()
775 instr1.AppendPartialToString(&str1); in TEST_F()
/external/swiftshader/third_party/llvm-7.0/llvm/docs/
DMergeFunctions.rst415 instr0 i32 %pf0 instr1 i32 %pf1 instr2 i32 123
421 instr0 i32 %pg0 instr1 i32 %pg0 instr2 i32 123
430 Instruction with opcode "*instr1*" from *f* is *greater* than instruction with
431 opcode "*instr1*" from *g*; here we have equal types and opcodes, but "*pf1* is
/external/llvm/docs/
DMergeFunctions.rst415 instr0 i32 %pf0 instr1 i32 %pf1 instr2 i32 123
421 instr0 i32 %pg0 instr1 i32 %pg0 instr2 i32 123
430 Instruction with opcode "*instr1*" from *f* is *greater* than instruction with
431 opcode "*instr1*" from *g*; here we have equal types and opcodes, but "*pf1* is
/external/v8/src/codegen/mips64/
Dassembler-mips64.h181 Instr instr1 = in pc_offset_for_safepoint() local
185 if (GetOpcodeField(instr1) != SPECIAL) { // instr1 == jialc. in pc_offset_for_safepoint()
186 DCHECK((kArchVariant == kMips64r6) && GetOpcodeField(instr1) == POP76 && in pc_offset_for_safepoint()
187 GetRs(instr1) == 0); in pc_offset_for_safepoint()
189 if (GetFunctionField(instr1) == SLL) { // instr1 == nop, instr2 == jalr. in pc_offset_for_safepoint()
193 DCHECK(GetFunctionField(instr1) == JALR); in pc_offset_for_safepoint()
Dassembler-mips64.cc3897 Instr instr1 = instr_at(pc + 1 * kInstrSize); in target_address_at() local
3902 if ((GetOpcodeField(instr0) == LUI) && (GetOpcodeField(instr1) == ORI) && in target_address_at()
3907 ((uint64_t)(GetImmediate16(instr1)) << 16) | in target_address_at()
3937 Instr instr1 = instr_at(pc + kInstrSize); in set_target_value_at() local
3938 uint32_t rt_code = GetRt(instr1); in set_target_value_at()
3945 DCHECK((GetOpcodeField(instr0) == LUI && GetOpcodeField(instr1) == ORI && in set_target_value_at()
/external/v8/src/compiler/backend/arm/
Dcode-generator-arm.cc406 #define ASSEMBLE_ATOMIC64_ARITH_BINOP(instr1, instr2) \ argument
413 __ instr1(i.TempRegister(1), r2, i.InputRegister(0), SBit::SetCC); \
3424 #define ATOMIC_ARITH_BINOP_CASE(op, instr1, instr2) \ in AssembleArchInstruction() argument
3427 ASSEMBLE_ATOMIC64_ARITH_BINOP(instr1, instr2); \ in AssembleArchInstruction()
3433 #define ATOMIC_LOGIC_BINOP_CASE(op, instr1) \ in AssembleArchInstruction() argument
3436 ASSEMBLE_ATOMIC64_LOGIC_BINOP(instr1); \ in AssembleArchInstruction()

12