| /external/llvm/lib/Target/WebAssembly/InstPrinter/ |
| WebAssemblyInstPrinter.cpp | 139 const MCOperand &Op = MI->getOperand(OpNo);
140 if (Op.isReg()) {
144 unsigned WAReg = Op.getReg();
156 } else if (Op.isImm()) {
166 O << Op.getImm();
167 } else if (Op.isFPImm()) {
177 O << toString(APFloat(float(Op.getFPImm())));
180 O << toString(APFloat(Op.getFPImm()));
188 assert(Op.isExpr() && "unknown operand kind in printOperand");
189 Op.getExpr()->print(O, &MAI)
[all...] |
| /external/spirv-llvm/lib/SPIRV/libSPIRV/ |
| SPIRVDecorate.h | 55 SPIRVDecorateGeneric(Op OC, SPIRVWord WC, Decoration TheDec,
58 SPIRVDecorateGeneric(Op OC, SPIRVWord WC, Decoration TheDec,
61 SPIRVDecorateGeneric(Op OC);
129 static const Op OC = OpDecorate;
199 static const Op OC = OpMemberDecorate;
234 static const Op OC = OpDecorationGroup;
269 SPIRVGroupDecorateGeneric(Op OC, SPIRVDecorationGroup *TheGroup,
276 SPIRVGroupDecorateGeneric(Op OC)
292 static const Op OC = OpGroupDecorate;
306 static const Op OC = OpGroupMemberDecorate
[all...] |
| /external/swiftshader/third_party/llvm-7.0/llvm/lib/Target/NVPTX/InstPrinter/ |
| NVPTXInstPrinter.cpp | 86 const MCOperand &Op = MI->getOperand(OpNo);
87 if (Op.isReg()) {
88 unsigned Reg = Op.getReg();
90 } else if (Op.isImm()) {
91 O << markup("<imm:") << formatImm(Op.getImm()) << markup(">");
93 assert(Op.isExpr() && "Unknown operand kind in printOperand");
94 Op.getExpr()->print(O, &MAI);
291 const MCOperand &Op = MI->getOperand(OpNum);
292 assert(Op.isExpr() && "Call prototype is not an MCExpr?");
293 const MCExpr *Expr = Op.getExpr()
[all...] |
| /external/swiftshader/third_party/llvm-7.0/llvm/lib/Target/WebAssembly/InstPrinter/ |
| WebAssemblyInstPrinter.cpp | 142 const MCOperand &Op = MI->getOperand(OpNo);
143 if (Op.isReg()) {
147 unsigned WAReg = Op.getReg();
159 } else if (Op.isImm()) {
170 O << Op.getImm();
171 } else if (Op.isFPImm()) {
181 O << ::toString(APFloat(float(Op.getFPImm())));
184 O << ::toString(APFloat(Op.getFPImm()));
192 assert(Op.isExpr() && "unknown operand kind in printOperand");
193 Op.getExpr()->print(O, &MAI)
[all...] |
| /external/tensorflow/tensorflow/compiler/xla/service/gpu/ |
| cudnn_fused_conv_rewriter.cc | 51 using match::Op;
63 if (!Match(instr, Maximum(zero_pattern, Op(&relu_input))) &&
64 !Match(instr, Maximum(Op(&relu_input), zero_pattern))) {
79 const auto bias_pattern = Broadcast(&bias_broadcast_instr, Op(&bias));
83 >e, Op(&conv_instr).WithOpcode(HloOpcode::kCustomCall), 0);
95 auto side_input_pattern = Op(&side_input);
112 auto add2_pattern = Add(&add1, Op(&addends[0]), Op(&addends[1]));
114 AddAnyOrder(&add2, add2_pattern, Op(&addends[2])), add2_pattern,
115 Op(&addends[0]))
[all...] |
| /external/swiftshader/third_party/LLVM/lib/Target/Sparc/ |
| SparcISelLowering.cpp | [all...] |
| /external/swiftshader/third_party/LLVM/lib/Target/X86/ |
| X86ISelLowering.cpp | [all...] |
| /external/llvm/lib/CodeGen/SelectionDAG/ |
| LegalizeTypes.h | 165 SDValue BitConvertToInteger(SDValue Op);
166 SDValue BitConvertVectorToIntegerVector(SDValue Op);
167 SDValue CreateStackStoreLoad(SDValue Op, EVT DestVT);
191 void SplitInteger(SDValue Op, SDValue &Lo, SDValue &Hi);
192 void SplitInteger(SDValue Op, EVT LoVT, EVT HiVT,
199 /// Given a processed operand Op which was promoted to a larger integer type,
201 /// corresponding to the original type are exactly equal to Op.
205 /// For example, if Op is an i16 and was promoted to an i32, then this method
206 /// returns an i32, the lower 16 bits of which coincide with Op, and the upper
208 SDValue GetPromotedInteger(SDValue Op) {
[all...] |
| /external/swiftshader/third_party/llvm-7.0/llvm/lib/CodeGen/SelectionDAG/ |
| LegalizeTypes.h | 216 SDValue BitConvertToInteger(SDValue Op);
217 SDValue BitConvertVectorToIntegerVector(SDValue Op);
218 SDValue CreateStackStoreLoad(SDValue Op, EVT DestVT);
237 void SplitInteger(SDValue Op, SDValue &Lo, SDValue &Hi);
238 void SplitInteger(SDValue Op, EVT LoVT, EVT HiVT,
250 /// Given a processed operand Op which was promoted to a larger integer type,
252 /// corresponding to the original type are exactly equal to Op.
256 /// For example, if Op is an i16 and was promoted to an i32, then this method
257 /// returns an i32, the lower 16 bits of which coincide with Op, and the upper
259 SDValue GetPromotedInteger(SDValue Op) {
[all...] |
| /external/llvm/lib/Target/Mips/ |
| MipsISelLowering.h | 250 SDValue LowerOperation(SDValue Op, SelectionDAG &DAG) const override;
393 SDValue lowerLOAD(SDValue Op, SelectionDAG &DAG) const;
394 SDValue lowerSTORE(SDValue Op, SelectionDAG &DAG) const;
431 SDValue lowerBR_JT(SDValue Op, SelectionDAG &DAG) const;
432 SDValue lowerBRCOND(SDValue Op, SelectionDAG &DAG) const;
433 SDValue lowerConstantPool(SDValue Op, SelectionDAG &DAG) const;
434 SDValue lowerGlobalAddress(SDValue Op, SelectionDAG &DAG) const;
435 SDValue lowerBlockAddress(SDValue Op, SelectionDAG &DAG) const;
436 SDValue lowerGlobalTLSAddress(SDValue Op, SelectionDAG &DAG) const;
437 SDValue lowerJumpTable(SDValue Op, SelectionDAG &DAG) const
[all...] |
| /external/swiftshader/third_party/llvm-7.0/llvm/lib/Target/X86/ |
| X86ISelLowering.cpp | [all...] |
| X86FixupSetCC.cpp | 105 for (auto &Op : MI->implicit_operands())
106 if ((Op.getReg() == X86::EFLAGS) && (Op.isDef()))
113 for (auto &Op : MI->implicit_operands())
114 if ((Op.getReg() == X86::EFLAGS) && (Op.isUse()))
|
| /external/swiftshader/third_party/LLVM/lib/Target/CellSPU/ |
| SPUISelLowering.cpp | 61 ExpandLibCall(RTLIB::Libcall LC, SDValue Op, SelectionDAG &DAG,
70 for (unsigned i = 0, e = Op.getNumOperands(); i != e; ++i) {
71 EVT ArgVT = Op.getOperand(i).getValueType();
73 Entry.Node = Op.getOperand(i);
84 Op.getNode()->getValueType(0).getTypeForEVT(*DAG.getContext());
89 Callee, Args, DAG, Op.getDebugLoc());
549 LowerLOAD(SDValue Op, SelectionDAG &DAG, const SPUSubtarget *ST) {
550 LoadSDNode *LN = cast<LoadSDNode>(Op);
554 EVT OutVT = Op.getValueType();
558 DebugLoc dl = Op.getDebugLoc()
[all...] |
| /external/llvm/lib/Target/ARM/InstPrinter/ |
| ARMInstPrinter.cpp | 274 const MCOperand &Op = MI->getOperand(OpNo);
275 if (Op.isReg()) {
276 unsigned Reg = Op.getReg();
278 } else if (Op.isImm()) {
279 O << markup("<imm:") << '#' << formatImm(Op.getImm()) << markup(">");
281 assert(Op.isExpr() && "unknown operand kind in printOperand");
282 const MCExpr *Expr = Op.getExpr();
379 void ARMInstPrinter::printAM2PreOrOffsetIndexOp(const MCInst *MI, unsigned Op,
382 const MCOperand &MO1 = MI->getOperand(Op);
383 const MCOperand &MO2 = MI->getOperand(Op + 1)
[all...] |
| /external/llvm/lib/Target/Hexagon/Disassembler/ |
| HexagonDisassembler.cpp | 101 static unsigned GetSubinstOpcode(unsigned IClass, unsigned inst, unsigned &op,
[all...] |
| /external/swiftshader/third_party/llvm-7.0/llvm/lib/Target/ARM/InstPrinter/ |
| ARMInstPrinter.cpp | 286 const MCOperand &Op = MI->getOperand(OpNo);
287 if (Op.isReg()) {
288 unsigned Reg = Op.getReg();
290 } else if (Op.isImm()) {
291 O << markup("<imm:") << '#' << formatImm(Op.getImm()) << markup(">");
293 assert(Op.isExpr() && "unknown operand kind in printOperand");
294 const MCExpr *Expr = Op.getExpr();
391 void ARMInstPrinter::printAM2PreOrOffsetIndexOp(const MCInst *MI, unsigned Op,
394 const MCOperand &MO1 = MI->getOperand(Op);
395 const MCOperand &MO2 = MI->getOperand(Op + 1)
[all...] |
| /external/tensorflow/tensorflow/core/kernels/ |
| reduction_gpu_kernels.cu.h | 58 // the overloaded addition op for std::complex
87 // the overloaded multiply op for std::complex
193 template <typename T, typename outT, int num_threads, typename Op>
195 T in, outT out, int num_elems, Op op,
209 sum = op(sum, in[pos]);
224 sum = BlockReduce(temp_storage).Reduce(sum, op, num_elements_to_reduce);
230 template <typename T, typename outT, typename Op>
232 T in, outT out, int num_rows, int num_cols, Op op,
[all...] |
| /external/llvm/lib/Target/X86/ |
| X86FixupSetCC.cpp | 104 for (auto &Op : MI->implicit_operands())
105 if ((Op.getReg() == X86::EFLAGS) && (Op.isDef()))
112 for (auto &Op : MI->implicit_operands())
113 if ((Op.getReg() == X86::EFLAGS) && (Op.isUse()))
|
| /external/llvm/utils/TableGen/ |
| DAGISelEmitter.cpp | 46 Record *Op = P->getOperator();
47 if (Op->isSubClassOf("Instruction")) {
49 CodeGenInstruction &II = CGP.getTargetInfo().getInstruction(Op);
65 Record *Op = P->getOperator();
66 if (Op->isSubClassOf("Instruction")) {
67 Cost += Op->getValueAsInt("CodeSize");
|
| /external/swiftshader/third_party/LLVM/lib/Target/SystemZ/ |
| SystemZISelLowering.h | 63 virtual SDValue LowerOperation(SDValue Op, SelectionDAG &DAG) const;
74 SDValue LowerBR_CC(SDValue Op, SelectionDAG &DAG) const;
75 SDValue LowerSELECT_CC(SDValue Op, SelectionDAG &DAG) const;
76 SDValue LowerGlobalAddress(SDValue Op, SelectionDAG &DAG) const;
77 SDValue LowerJumpTable(SDValue Op, SelectionDAG &DAG) const;
78 SDValue LowerConstantPool(SDValue Op, SelectionDAG &DAG) const;
|
| /external/swiftshader/third_party/LLVM/utils/TableGen/ |
| DAGISelEmitter.cpp | 32 Record *Op = P->getOperator();
33 if (Op->isSubClassOf("Instruction")) {
35 CodeGenInstruction &II = CGP.getTargetInfo().getInstruction(Op);
51 Record *Op = P->getOperator();
52 if (Op->isSubClassOf("Instruction")) {
53 Cost += Op->getValueAsInt("CodeSize");
|
| /external/swiftshader/third_party/llvm-7.0/llvm/lib/Target/Mips/ |
| MipsISelLowering.h | 322 SDValue LowerOperation(SDValue Op, SelectionDAG &DAG) const override;
502 SDValue lowerLOAD(SDValue Op, SelectionDAG &DAG) const;
503 SDValue lowerSTORE(SDValue Op, SelectionDAG &DAG) const;
540 SDValue lowerBRCOND(SDValue Op, SelectionDAG &DAG) const;
541 SDValue lowerConstantPool(SDValue Op, SelectionDAG &DAG) const;
542 SDValue lowerGlobalAddress(SDValue Op, SelectionDAG &DAG) const;
543 SDValue lowerBlockAddress(SDValue Op, SelectionDAG &DAG) const;
544 SDValue lowerGlobalTLSAddress(SDValue Op, SelectionDAG &DAG) const;
545 SDValue lowerJumpTable(SDValue Op, SelectionDAG &DAG) const;
546 SDValue lowerSELECT(SDValue Op, SelectionDAG &DAG) const
[all...] |
| /external/swiftshader/third_party/llvm-7.0/llvm/utils/TableGen/ |
| DAGISelEmitter.cpp | 46 Record *Op = P->getOperator();
47 if (Op->isSubClassOf("Instruction")) {
49 CodeGenInstruction &II = CGP.getTargetInfo().getInstruction(Op);
65 Record *Op = P->getOperator();
66 if (Op->isSubClassOf("Instruction")) {
67 Cost += Op->getValueAsInt("CodeSize");
|
| /external/llvm/lib/Target/AMDGPU/ |
| AMDGPUISelLowering.cpp | 687 SDValue AMDGPUTargetLowering::LowerDYNAMIC_STACKALLOC(SDValue Op,
692 SDLoc(Op).getDebugLoc());
694 auto Ops = {DAG.getConstant(0, SDLoc(), Op.getValueType()), Op.getOperand(0)};
698 SDValue AMDGPUTargetLowering::LowerOperation(SDValue Op,
700 switch (Op.getOpcode()) {
702 Op->dump(&DAG);
706 case ISD::SIGN_EXTEND_INREG: return LowerSIGN_EXTEND_INREG(Op, DAG);
707 case ISD::CONCAT_VECTORS: return LowerCONCAT_VECTORS(Op, DAG);
708 case ISD::EXTRACT_SUBVECTOR: return LowerEXTRACT_SUBVECTOR(Op, DAG)
[all...] |
| /external/swiftshader/third_party/LLVM/lib/Target/MSP430/ |
| MSP430ISelLowering.cpp | 179 SDValue MSP430TargetLowering::LowerOperation(SDValue Op,
181 switch (Op.getOpcode()) {
184 case ISD::SRA: return LowerShifts(Op, DAG);
185 case ISD::GlobalAddress: return LowerGlobalAddress(Op, DAG);
186 case ISD::BlockAddress: return LowerBlockAddress(Op, DAG);
187 case ISD::ExternalSymbol: return LowerExternalSymbol(Op, DAG);
188 case ISD::SETCC: return LowerSETCC(Op, DAG);
189 case ISD::BR_CC: return LowerBR_CC(Op, DAG);
190 case ISD::SELECT_CC: return LowerSELECT_CC(Op, DAG);
191 case ISD::SIGN_EXTEND: return LowerSIGN_EXTEND(Op, DAG)
[all...] |
