[mlir] Move casting calls from methods to function calls

[tpopp]

Summary

The MLIR classes Type/Attribute/Operation/Op/Value support
cast/dyn_cast/isa/dyn_cast_or_null functionality through llvm's doCast
functionality in addition to defining methods with the same name.
This change begins the migration of uses of the method to the
corresponding function call as has been decided as more consistent.

Note that there still exist classes that only define methods directly,
such as AffineExpr, and this does not include work currently to support
a functional cast/isa call.

This prefixes everything with llvm:: because some places require it, it would take more time to figure
out which subset of files don't need it, and they can be removed easily once the methods are removed.

Context:

• https://mlir.llvm.org/deprecation/ at "Use the free function variants for dyn_cast/cast/isa/…"
• Original discussion at https://discourse.llvm.org/t/preferred-casting-style-going-forward/68443

Implementation:
This patch updates all remaining uses of the deprecated functionality in
mlir/. This was done with clang-tidy as described below and further
modifications to GPUBase.td and OpenMPOpsInterfaces.td.

Steps are described per line, as comments are removed by git:
0. Retrieve the change from the following to build clang-tidy with an

additional check:
main...tpopp:llvm-project:tidy-cast-check

1. Build clang-tidy
2. Run clang-tidy over your entire codebase while disabling all checks and enabling the one relevant one. Run on all header files also.
3. Delete .inc files that were also modified, so the next build rebuilds them to a pure state.

ninja -C $BUILD_DIR clang-tidy

run-clang-tidy -clang-tidy-binary=$BUILD_DIR/bin/clang-tidy -checks='-*,misc-cast-functions'\
               -header-filter=mlir/ mlir/* -fix

rm -rf $BUILD_DIR/tools/mlir/**/*.inc

[mehdi_amini]

mlir/include/mlir/Debug/BreakpointManagers/FileLineColLocBreakpointManager.h

77

Why "dyn_cast_if_present"? These should all be non-null and dyn_cast should be enough?

mlir/examples/toy/Ch2/mlir/Dialect.cpp

static mlir::ParseResult parseBinaryOp(mlir::OpAsmParser &parser,
   Type type;
   if (parser.parseOperandList(operands, /*requiredOperandCount=*/2) ||
       parser.parseOptionalAttrDict(result.attributes) ||
       parser.parseColonType(type))
     return mlir::failure();
 
   // If the type is a function type, it contains the input and result types of
   // this operation.
-  if (FunctionType funcType = type.dyn_cast<FunctionType>()) {
+  if (FunctionType funcType = llvm::dyn_cast<FunctionType>(type)) {
     if (parser.resolveOperands(operands, funcType.getInputs(), operandsLoc,
                                result.operands))
       return mlir::failure();
     result.addTypes(funcType.getResults());
     return mlir::success();
   }
 
   // Otherwise, the parsed type is the type of both operands and results.
void ConstantOp::print(mlir::OpAsmPrinter &printer) {
   printer << getValue();
 }
 
 /// Verifier for the constant operation. This corresponds to the
 /// `let hasVerifier = 1` in the op definition.
 mlir::LogicalResult ConstantOp::verify() {
   // If the return type of the constant is not an unranked tensor, the shape
   // must match the shape of the attribute holding the data.
-  auto resultType = getResult().getType().dyn_cast<mlir::RankedTensorType>();
+  auto resultType = llvm::dyn_cast<mlir::RankedTensorType>(getResult().getType());
   if (!resultType)
     return success();
 
   // Check that the rank of the attribute type matches the rank of the constant
   // result type.
-  auto attrType = getValue().getType().cast<mlir::RankedTensorType>();
+  auto attrType = llvm::cast<mlir::RankedTensorType>(getValue().getType());
   if (attrType.getRank() != resultType.getRank()) {
     return emitOpError("return type must match the one of the attached value "
                        "attribute: ")
            << attrType.getRank() << " != " << resultType.getRank();
   }
 
   // Check that each of the dimensions match between the two types.
   for (int dim = 0, dimE = attrType.getRank(); dim < dimE; ++dim) {
mlir::LogicalResult ReturnOp::verify() {
   // If the operation does not have an input, we are done.
   if (!hasOperand())
     return mlir::success();
 
   auto inputType = *operand_type_begin();
   auto resultType = results.front();
 
   // Check that the result type of the function matches the operand type.
-  if (inputType == resultType || inputType.isa<mlir::UnrankedTensorType>() ||
-      resultType.isa<mlir::UnrankedTensorType>())
+  if (inputType == resultType || llvm::isa<mlir::UnrankedTensorType>(inputType) ||
+      llvm::isa<mlir::UnrankedTensorType>(resultType))
     return mlir::success();
 
   return emitError() << "type of return operand (" << inputType
                      << ") doesn't match function result type (" << resultType
                      << ")";
 }
 
 //===----------------------------------------------------------------------===//
 // TransposeOp
 //===----------------------------------------------------------------------===//
 
 void TransposeOp::build(mlir::OpBuilder &builder, mlir::OperationState &state,
                         mlir::Value value) {
   state.addTypes(UnrankedTensorType::get(builder.getF64Type()));
   state.addOperands(value);
 }
 
 mlir::LogicalResult TransposeOp::verify() {
-  auto inputType = getOperand().getType().dyn_cast<RankedTensorType>();
-  auto resultType = getType().dyn_cast<RankedTensorType>();
+  auto inputType = llvm::dyn_cast<RankedTensorType>(getOperand().getType());
+  auto resultType = llvm::dyn_cast<RankedTensorType>(getType());
   if (!inputType || !resultType)
     return mlir::success();
 
   auto inputShape = inputType.getShape();
   if (!std::equal(inputShape.begin(), inputShape.end(),
                   resultType.getShape().rbegin())) {
     return emitError()
            << "expected result shape to be a transpose of the input";

mlir/examples/toy/Ch3/mlir/Dialect.cpp

static mlir::ParseResult parseBinaryOp(mlir::OpAsmParser &parser,
   Type type;
   if (parser.parseOperandList(operands, /*requiredOperandCount=*/2) ||
       parser.parseOptionalAttrDict(result.attributes) ||
       parser.parseColonType(type))
     return mlir::failure();
 
   // If the type is a function type, it contains the input and result types of
   // this operation.
-  if (FunctionType funcType = type.dyn_cast<FunctionType>()) {
+  if (FunctionType funcType = llvm::dyn_cast<FunctionType>(type)) {
     if (parser.resolveOperands(operands, funcType.getInputs(), operandsLoc,
                                result.operands))
       return mlir::failure();
     result.addTypes(funcType.getResults());
     return mlir::success();
   }
 
   // Otherwise, the parsed type is the type of both operands and results.
void ConstantOp::print(mlir::OpAsmPrinter &printer) {
   printer << getValue();
 }
 
 /// Verifier for the constant operation. This corresponds to the
 /// `let hasVerifier = 1` in the op definition.
 mlir::LogicalResult ConstantOp::verify() {
   // If the return type of the constant is not an unranked tensor, the shape
   // must match the shape of the attribute holding the data.
-  auto resultType = getResult().getType().dyn_cast<mlir::RankedTensorType>();
+  auto resultType = llvm::dyn_cast<mlir::RankedTensorType>(getResult().getType());
   if (!resultType)
     return success();
 
   // Check that the rank of the attribute type matches the rank of the constant
   // result type.
-  auto attrType = getValue().getType().cast<mlir::RankedTensorType>();
+  auto attrType = llvm::cast<mlir::RankedTensorType>(getValue().getType());
   if (attrType.getRank() != resultType.getRank()) {
     return emitOpError("return type must match the one of the attached value "
                        "attribute: ")
            << attrType.getRank() << " != " << resultType.getRank();
   }
 
   // Check that each of the dimensions match between the two types.
   for (int dim = 0, dimE = attrType.getRank(); dim < dimE; ++dim) {
mlir::LogicalResult ReturnOp::verify() {
   // If the operation does not have an input, we are done.
   if (!hasOperand())
     return mlir::success();
 
   auto inputType = *operand_type_begin();
   auto resultType = results.front();
 
   // Check that the result type of the function matches the operand type.
-  if (inputType == resultType || inputType.isa<mlir::UnrankedTensorType>() ||
-      resultType.isa<mlir::UnrankedTensorType>())
+  if (inputType == resultType || llvm::isa<mlir::UnrankedTensorType>(inputType) ||
+      llvm::isa<mlir::UnrankedTensorType>(resultType))
     return mlir::success();
 
   return emitError() << "type of return operand (" << inputType
                      << ") doesn't match function result type (" << resultType
                      << ")";
 }
 
 //===----------------------------------------------------------------------===//
 // TransposeOp
 //===----------------------------------------------------------------------===//
 
 void TransposeOp::build(mlir::OpBuilder &builder, mlir::OperationState &state,
                         mlir::Value value) {
   state.addTypes(UnrankedTensorType::get(builder.getF64Type()));
   state.addOperands(value);
 }
 
 mlir::LogicalResult TransposeOp::verify() {
-  auto inputType = getOperand().getType().dyn_cast<RankedTensorType>();
-  auto resultType = getType().dyn_cast<RankedTensorType>();
+  auto inputType = llvm::dyn_cast<RankedTensorType>(getOperand().getType());
+  auto resultType = llvm::dyn_cast<RankedTensorType>(getType());
   if (!inputType || !resultType)
     return mlir::success();
 
   auto inputShape = inputType.getShape();
   if (!std::equal(inputShape.begin(), inputShape.end(),
                   resultType.getShape().rbegin())) {
     return emitError()
            << "expected result shape to be a transpose of the input";

mlir/examples/toy/Ch4/mlir/Dialect.cpp

static mlir::ParseResult parseBinaryOp(mlir::OpAsmParser &parser,
   Type type;
   if (parser.parseOperandList(operands, /*requiredOperandCount=*/2) ||
       parser.parseOptionalAttrDict(result.attributes) ||
       parser.parseColonType(type))
     return mlir::failure();
 
   // If the type is a function type, it contains the input and result types of
   // this operation.
-  if (FunctionType funcType = type.dyn_cast<FunctionType>()) {
+  if (FunctionType funcType = llvm::dyn_cast<FunctionType>(type)) {
     if (parser.resolveOperands(operands, funcType.getInputs(), operandsLoc,
                                result.operands))
       return mlir::failure();
     result.addTypes(funcType.getResults());
     return mlir::success();
   }
 
   // Otherwise, the parsed type is the type of both operands and results.
void ConstantOp::print(mlir::OpAsmPrinter &printer) {
   printer << getValue();
 }
 
 /// Verifier for the constant operation. This corresponds to the
 /// `let hasVerifier = 1` in the op definition.
 mlir::LogicalResult ConstantOp::verify() {
   // If the return type of the constant is not an unranked tensor, the shape
   // must match the shape of the attribute holding the data.
-  auto resultType = getResult().getType().dyn_cast<mlir::RankedTensorType>();
+  auto resultType = llvm::dyn_cast<mlir::RankedTensorType>(getResult().getType());
   if (!resultType)
     return success();
 
   // Check that the rank of the attribute type matches the rank of the constant
   // result type.
-  auto attrType = getValue().getType().cast<mlir::RankedTensorType>();
+  auto attrType = llvm::cast<mlir::RankedTensorType>(getValue().getType());
   if (attrType.getRank() != resultType.getRank()) {
     return emitOpError("return type must match the one of the attached value "
                        "attribute: ")
            << attrType.getRank() << " != " << resultType.getRank();
   }
 
   // Check that each of the dimensions match between the two types.
   for (int dim = 0, dimE = attrType.getRank(); dim < dimE; ++dim) {

 
 /// Returns true if the given set of input and result types are compatible with
 /// this cast operation. This is required by the `CastOpInterface` to verify
 /// this operation and provide other additional utilities.
 bool CastOp::areCastCompatible(TypeRange inputs, TypeRange outputs) {
   if (inputs.size() != 1 || outputs.size() != 1)
     return false;
   // The inputs must be Tensors with the same element type.
-  TensorType input = inputs.front().dyn_cast<TensorType>();
-  TensorType output = outputs.front().dyn_cast<TensorType>();
+  TensorType input = llvm::dyn_cast<TensorType>(inputs.front());
+  TensorType output = llvm::dyn_cast<TensorType>(outputs.front());
   if (!input || !output || input.getElementType() != output.getElementType())
     return false;
   // The shape is required to match if both types are ranked.
   return !input.hasRank() || !output.hasRank() || input == output;
 }
 
 //===----------------------------------------------------------------------===//
 // FuncOp
mlir::LogicalResult ReturnOp::verify() {
   // If the operation does not have an input, we are done.
   if (!hasOperand())
     return mlir::success();
 
   auto inputType = *operand_type_begin();
   auto resultType = results.front();
 
   // Check that the result type of the function matches the operand type.
-  if (inputType == resultType || inputType.isa<mlir::UnrankedTensorType>() ||
-      resultType.isa<mlir::UnrankedTensorType>())
+  if (inputType == resultType || llvm::isa<mlir::UnrankedTensorType>(inputType) ||
+      llvm::isa<mlir::UnrankedTensorType>(resultType))
     return mlir::success();
 
   return emitError() << "type of return operand (" << inputType
                      << ") doesn't match function result type (" << resultType
                      << ")";
 }
 
 //===----------------------------------------------------------------------===//
 // TransposeOp
 //===----------------------------------------------------------------------===//
 
 void TransposeOp::build(mlir::OpBuilder &builder, mlir::OperationState &state,
                         mlir::Value value) {
   state.addTypes(UnrankedTensorType::get(builder.getF64Type()));
   state.addOperands(value);
 }
 
 void TransposeOp::inferShapes() {
-  auto arrayTy = getOperand().getType().cast<RankedTensorType>();
+  auto arrayTy = llvm::cast<RankedTensorType>(getOperand().getType());
   SmallVector<int64_t, 2> dims(llvm::reverse(arrayTy.getShape()));
   getResult().setType(RankedTensorType::get(dims, arrayTy.getElementType()));
 }
 
 mlir::LogicalResult TransposeOp::verify() {
-  auto inputType = getOperand().getType().dyn_cast<RankedTensorType>();
-  auto resultType = getType().dyn_cast<RankedTensorType>();
+  auto inputType = llvm::dyn_cast<RankedTensorType>(getOperand().getType());
+  auto resultType = llvm::dyn_cast<RankedTensorType>(getType());
   if (!inputType || !resultType)
     return mlir::success();
 
   auto inputShape = inputType.getShape();
   if (!std::equal(inputShape.begin(), inputShape.end(),
                   resultType.getShape().rbegin())) {
     return emitError()
            << "expected result shape to be a transpose of the input";

mlir/examples/toy/Ch4/mlir/ShapeInferencePass.cpp

if (!opWorklist.empty()) {
       signalPassFailure();
     }
   }
 
   /// A utility method that returns if the given operation has all of its
   /// operands inferred.
   static bool allOperandsInferred(Operation *op) {
     return llvm::all_of(op->getOperandTypes(), [](Type operandType) {
-      return operandType.isa<RankedTensorType>();
+      return llvm::isa<RankedTensorType>(operandType);
     });
   }
 
   /// A utility method that returns if the given operation has a dynamically
   /// shaped result.
   static bool returnsDynamicShape(Operation *op) {
     return llvm::any_of(op->getResultTypes(), [](Type resultType) {
-      return !resultType.isa<RankedTensorType>();
+      return !llvm::isa<RankedTensorType>(resultType);
     });
   }
 };
 } // namespace
 
 /// Create a Shape Inference pass.
 std::unique_ptr<mlir::Pass> mlir::toy::createShapeInferencePass() {
   return std::make_unique<ShapeInferencePass>();
 }

mlir/examples/toy/Ch5/mlir/Dialect.cpp

static mlir::ParseResult parseBinaryOp(mlir::OpAsmParser &parser,
   Type type;
   if (parser.parseOperandList(operands, /*requiredOperandCount=*/2) ||
       parser.parseOptionalAttrDict(result.attributes) ||
       parser.parseColonType(type))
     return mlir::failure();
 
   // If the type is a function type, it contains the input and result types of
   // this operation.
-  if (FunctionType funcType = type.dyn_cast<FunctionType>()) {
+  if (FunctionType funcType = llvm::dyn_cast<FunctionType>(type)) {
     if (parser.resolveOperands(operands, funcType.getInputs(), operandsLoc,
                                result.operands))
       return mlir::failure();
     result.addTypes(funcType.getResults());
     return mlir::success();
   }
 
   // Otherwise, the parsed type is the type of both operands and results.
void ConstantOp::print(mlir::OpAsmPrinter &printer) {
   printer << getValue();
 }
 
 /// Verifier for the constant operation. This corresponds to the
 /// `let hasVerifier = 1` in the op definition.
 mlir::LogicalResult ConstantOp::verify() {
   // If the return type of the constant is not an unranked tensor, the shape
   // must match the shape of the attribute holding the data.
-  auto resultType = getResult().getType().dyn_cast<mlir::RankedTensorType>();
+  auto resultType = llvm::dyn_cast<mlir::RankedTensorType>(getResult().getType());
   if (!resultType)
     return success();
 
   // Check that the rank of the attribute type matches the rank of the constant
   // result type.
-  auto attrType = getValue().getType().cast<mlir::RankedTensorType>();
+  auto attrType = llvm::cast<mlir::RankedTensorType>(getValue().getType());
   if (attrType.getRank() != resultType.getRank()) {
     return emitOpError("return type must match the one of the attached value "
                        "attribute: ")
            << attrType.getRank() << " != " << resultType.getRank();
   }
 
   // Check that each of the dimensions match between the two types.
   for (int dim = 0, dimE = attrType.getRank(); dim < dimE; ++dim) {

 
 /// Returns true if the given set of input and result types are compatible with
 /// this cast operation. This is required by the `CastOpInterface` to verify
 /// this operation and provide other additional utilities.
 bool CastOp::areCastCompatible(TypeRange inputs, TypeRange outputs) {
   if (inputs.size() != 1 || outputs.size() != 1)
     return false;
   // The inputs must be Tensors with the same element type.
-  TensorType input = inputs.front().dyn_cast<TensorType>();
-  TensorType output = outputs.front().dyn_cast<TensorType>();
+  TensorType input = llvm::dyn_cast<TensorType>(inputs.front());
+  TensorType output = llvm::dyn_cast<TensorType>(outputs.front());
   if (!input || !output || input.getElementType() != output.getElementType())
     return false;
   // The shape is required to match if both types are ranked.
   return !input.hasRank() || !output.hasRank() || input == output;
 }
 
 //===----------------------------------------------------------------------===//
 // FuncOp
mlir::LogicalResult ReturnOp::verify() {
   // If the operation does not have an input, we are done.
   if (!hasOperand())
     return mlir::success();
 
   auto inputType = *operand_type_begin();
   auto resultType = results.front();
 
   // Check that the result type of the function matches the operand type.
-  if (inputType == resultType || inputType.isa<mlir::UnrankedTensorType>() ||
-      resultType.isa<mlir::UnrankedTensorType>())
+  if (inputType == resultType || llvm::isa<mlir::UnrankedTensorType>(inputType) ||
+      llvm::isa<mlir::UnrankedTensorType>(resultType))
     return mlir::success();
 
   return emitError() << "type of return operand (" << inputType
                      << ") doesn't match function result type (" << resultType
                      << ")";
 }
 
 //===----------------------------------------------------------------------===//
 // TransposeOp
 //===----------------------------------------------------------------------===//
 
 void TransposeOp::build(mlir::OpBuilder &builder, mlir::OperationState &state,
                         mlir::Value value) {
   state.addTypes(UnrankedTensorType::get(builder.getF64Type()));
   state.addOperands(value);
 }
 
 void TransposeOp::inferShapes() {
-  auto arrayTy = getOperand().getType().cast<RankedTensorType>();
+  auto arrayTy = llvm::cast<RankedTensorType>(getOperand().getType());
   SmallVector<int64_t, 2> dims(llvm::reverse(arrayTy.getShape()));
   getResult().setType(RankedTensorType::get(dims, arrayTy.getElementType()));
 }
 
 mlir::LogicalResult TransposeOp::verify() {
-  auto inputType = getOperand().getType().dyn_cast<RankedTensorType>();
-  auto resultType = getType().dyn_cast<RankedTensorType>();
+  auto inputType = llvm::dyn_cast<RankedTensorType>(getOperand().getType());
+  auto resultType = llvm::dyn_cast<RankedTensorType>(getType());
   if (!inputType || !resultType)
     return mlir::success();
 
   auto inputShape = inputType.getShape();
   if (!std::equal(inputShape.begin(), inputShape.end(),
                   resultType.getShape().rbegin())) {
     return emitError()
            << "expected result shape to be a transpose of the input";

mlir/examples/toy/Ch5/mlir/LowerToAffineLoops.cpp

 /// induction variables for the iteration. It returns a value to store at the
 /// current index of the iteration.
 using LoopIterationFn = function_ref<Value(
     OpBuilder &rewriter, ValueRange memRefOperands, ValueRange loopIvs)>;
 
 static void lowerOpToLoops(Operation *op, ValueRange operands,
                            PatternRewriter &rewriter,
                            LoopIterationFn processIteration) {
-  auto tensorType = (*op->result_type_begin()).cast<RankedTensorType>();
+  auto tensorType = llvm::cast<RankedTensorType>((*op->result_type_begin()));
   auto loc = op->getLoc();
 
   // Insert an allocation and deallocation for the result of this operation.
   auto memRefType = convertTensorToMemRef(tensorType);
   auto alloc = insertAllocAndDealloc(memRefType, loc, rewriter);
 
   // Create a nest of affine loops, with one loop per dimension of the shape.
   // The buildAffineLoopNest function takes a callback that is used to construct
struct ConstantOpLowering : public OpRewritePattern<toy::ConstantOp> {
 
   LogicalResult matchAndRewrite(toy::ConstantOp op,
                                 PatternRewriter &rewriter) const final {
     DenseElementsAttr constantValue = op.getValue();
     Location loc = op.getLoc();
 
     // When lowering the constant operation, we allocate and assign the constant
     // values to a corresponding memref allocation.
-    auto tensorType = op.getType().cast<RankedTensorType>();
+    auto tensorType = llvm::cast<RankedTensorType>(op.getType());
     auto memRefType = convertTensorToMemRef(tensorType);
     auto alloc = insertAllocAndDealloc(memRefType, loc, rewriter);
 
     // We will be generating constant indices up-to the largest dimension.
     // Create these constants up-front to avoid large amounts of redundant
     // operations.
     auto valueShape = memRefType.getShape();
     SmallVector<Value, 8> constantIndices;
void ToyToAffineLoweringPass::runOnOperation() {
   // if any of these operations are *not* converted. Given that we actually want
   // a partial lowering, we explicitly mark the Toy operations that don't want
   // to lower, `toy.print`, as `legal`. `toy.print` will still need its operands
   // to be updated though (as we convert from TensorType to MemRefType), so we
   // only treat it as `legal` if its operands are legal.
   target.addIllegalDialect<toy::ToyDialect>();
   target.addDynamicallyLegalOp<toy::PrintOp>([](toy::PrintOp op) {
     return llvm::none_of(op->getOperandTypes(),
-                         [](Type type) { return type.isa<TensorType>(); });
+                         [](Type type) { return llvm::isa<TensorType>(type); });
   });
 
   // Now that the conversion target has been defined, we just need to provide
   // the set of patterns that will lower the Toy operations.
   RewritePatternSet patterns(&getContext());
   patterns.add<AddOpLowering, ConstantOpLowering, FuncOpLowering, MulOpLowering,
                PrintOpLowering, ReturnOpLowering, TransposeOpLowering>(
       &getContext());

mlir/examples/toy/Ch5/mlir/ShapeInferencePass.cpp

if (!opWorklist.empty()) {
       signalPassFailure();
     }
   }
 
   /// A utility method that returns if the given operation has all of its
   /// operands inferred.
   static bool allOperandsInferred(Operation *op) {
     return llvm::all_of(op->getOperandTypes(), [](Type operandType) {
-      return operandType.isa<RankedTensorType>();
+      return llvm::isa<RankedTensorType>(operandType);
     });
   }
 
   /// A utility method that returns if the given operation has a dynamically
   /// shaped result.
   static bool returnsDynamicShape(Operation *op) {
     return llvm::any_of(op->getResultTypes(), [](Type resultType) {
-      return !resultType.isa<RankedTensorType>();
+      return !llvm::isa<RankedTensorType>(resultType);
     });
   }
 };
 } // namespace
 
 /// Create a Shape Inference pass.
 std::unique_ptr<mlir::Pass> mlir::toy::createShapeInferencePass() {
   return std::make_unique<ShapeInferencePass>();
 }

mlir/examples/toy/Ch6/mlir/Dialect.cpp

static mlir::ParseResult parseBinaryOp(mlir::OpAsmParser &parser,
   Type type;
   if (parser.parseOperandList(operands, /*requiredOperandCount=*/2) ||
       parser.parseOptionalAttrDict(result.attributes) ||
       parser.parseColonType(type))
     return mlir::failure();
 
   // If the type is a function type, it contains the input and result types of
   // this operation.
-  if (FunctionType funcType = type.dyn_cast<FunctionType>()) {
+  if (FunctionType funcType = llvm::dyn_cast<FunctionType>(type)) {
     if (parser.resolveOperands(operands, funcType.getInputs(), operandsLoc,
                                result.operands))
       return mlir::failure();
     result.addTypes(funcType.getResults());
     return mlir::success();
   }
 
   // Otherwise, the parsed type is the type of both operands and results.
void ConstantOp::print(mlir::OpAsmPrinter &printer) {
   printer << getValue();
 }
 
 /// Verifier for the constant operation. This corresponds to the
 /// `let hasVerifier = 1` in the op definition.
 mlir::LogicalResult ConstantOp::verify() {
   // If the return type of the constant is not an unranked tensor, the shape
   // must match the shape of the attribute holding the data.
-  auto resultType = getResult().getType().dyn_cast<mlir::RankedTensorType>();
+  auto resultType = llvm::dyn_cast<mlir::RankedTensorType>(getResult().getType());
   if (!resultType)
     return success();
 
   // Check that the rank of the attribute type matches the rank of the constant
   // result type.
-  auto attrType = getValue().getType().cast<mlir::RankedTensorType>();
+  auto attrType = llvm::cast<mlir::RankedTensorType>(getValue().getType());
   if (attrType.getRank() != resultType.getRank()) {
     return emitOpError("return type must match the one of the attached value "
                        "attribute: ")
            << attrType.getRank() << " != " << resultType.getRank();
   }
 
   // Check that each of the dimensions match between the two types.
   for (int dim = 0, dimE = attrType.getRank(); dim < dimE; ++dim) {

 
 /// Returns true if the given set of input and result types are compatible with
 /// this cast operation. This is required by the `CastOpInterface` to verify
 /// this operation and provide other additional utilities.
 bool CastOp::areCastCompatible(TypeRange inputs, TypeRange outputs) {
   if (inputs.size() != 1 || outputs.size() != 1)
     return false;
   // The inputs must be Tensors with the same element type.
-  TensorType input = inputs.front().dyn_cast<TensorType>();
-  TensorType output = outputs.front().dyn_cast<TensorType>();
+  TensorType input = llvm::dyn_cast<TensorType>(inputs.front());
+  TensorType output = llvm::dyn_cast<TensorType>(outputs.front());
   if (!input || !output || input.getElementType() != output.getElementType())
     return false;
   // The shape is required to match if both types are ranked.
   return !input.hasRank() || !output.hasRank() || input == output;
 }
 
 //===----------------------------------------------------------------------===//
 // FuncOp
mlir::LogicalResult ReturnOp::verify() {
   // If the operation does not have an input, we are done.
   if (!hasOperand())
     return mlir::success();
 
   auto inputType = *operand_type_begin();
   auto resultType = results.front();
 
   // Check that the result type of the function matches the operand type.
-  if (inputType == resultType || inputType.isa<mlir::UnrankedTensorType>() ||
-      resultType.isa<mlir::UnrankedTensorType>())
+  if (inputType == resultType || llvm::isa<mlir::UnrankedTensorType>(inputType) ||
+      llvm::isa<mlir::UnrankedTensorType>(resultType))
     return mlir::success();
 
   return emitError() << "type of return operand (" << inputType
                      << ") doesn't match function result type (" << resultType
                      << ")";
 }
 
 //===----------------------------------------------------------------------===//
 // TransposeOp
 //===----------------------------------------------------------------------===//
 
 void TransposeOp::build(mlir::OpBuilder &builder, mlir::OperationState &state,
                         mlir::Value value) {
   state.addTypes(UnrankedTensorType::get(builder.getF64Type()));
   state.addOperands(value);
 }
 
 void TransposeOp::inferShapes() {
-  auto arrayTy = getOperand().getType().cast<RankedTensorType>();
+  auto arrayTy = llvm::cast<RankedTensorType>(getOperand().getType());
   SmallVector<int64_t, 2> dims(llvm::reverse(arrayTy.getShape()));
   getResult().setType(RankedTensorType::get(dims, arrayTy.getElementType()));
 }
 
 mlir::LogicalResult TransposeOp::verify() {
-  auto inputType = getOperand().getType().dyn_cast<RankedTensorType>();
-  auto resultType = getType().dyn_cast<RankedTensorType>();
+  auto inputType = llvm::dyn_cast<RankedTensorType>(getOperand().getType());
+  auto resultType = llvm::dyn_cast<RankedTensorType>(getType());
   if (!inputType || !resultType)
     return mlir::success();
 
   auto inputShape = inputType.getShape();
   if (!std::equal(inputShape.begin(), inputShape.end(),
                   resultType.getShape().rbegin())) {
     return emitError()
            << "expected result shape to be a transpose of the input";

mlir/examples/toy/Ch6/mlir/LowerToAffineLoops.cpp

 /// induction variables for the iteration. It returns a value to store at the
 /// current index of the iteration.
 using LoopIterationFn = function_ref<Value(
     OpBuilder &rewriter, ValueRange memRefOperands, ValueRange loopIvs)>;
 
 static void lowerOpToLoops(Operation *op, ValueRange operands,
                            PatternRewriter &rewriter,
                            LoopIterationFn processIteration) {
-  auto tensorType = (*op->result_type_begin()).cast<RankedTensorType>();
+  auto tensorType = llvm::cast<RankedTensorType>((*op->result_type_begin()));
   auto loc = op->getLoc();
 
   // Insert an allocation and deallocation for the result of this operation.
   auto memRefType = convertTensorToMemRef(tensorType);
   auto alloc = insertAllocAndDealloc(memRefType, loc, rewriter);
 
   // Create a nest of affine loops, with one loop per dimension of the shape.
   // The buildAffineLoopNest function takes a callback that is used to construct
struct ConstantOpLowering : public OpRewritePattern<toy::ConstantOp> {
 
   LogicalResult matchAndRewrite(toy::ConstantOp op,
                                 PatternRewriter &rewriter) const final {
     DenseElementsAttr constantValue = op.getValue();
     Location loc = op.getLoc();
 
     // When lowering the constant operation, we allocate and assign the constant
     // values to a corresponding memref allocation.
-    auto tensorType = op.getType().cast<RankedTensorType>();
+    auto tensorType = llvm::cast<RankedTensorType>(op.getType());
     auto memRefType = convertTensorToMemRef(tensorType);
     auto alloc = insertAllocAndDealloc(memRefType, loc, rewriter);
 
     // We will be generating constant indices up-to the largest dimension.
     // Create these constants up-front to avoid large amounts of redundant
     // operations.
     auto valueShape = memRefType.getShape();
     SmallVector<Value, 8> constantIndices;
void ToyToAffineLoweringPass::runOnOperation() {
   // if any of these operations are *not* converted. Given that we actually want
   // a partial lowering, we explicitly mark the Toy operations that don't want
   // to lower, `toy.print`, as `legal`. `toy.print` will still need its operands
   // to be updated though (as we convert from TensorType to MemRefType), so we
   // only treat it as `legal` if its operands are legal.
   target.addIllegalDialect<toy::ToyDialect>();
   target.addDynamicallyLegalOp<toy::PrintOp>([](toy::PrintOp op) {
     return llvm::none_of(op->getOperandTypes(),
-                         [](Type type) { return type.isa<TensorType>(); });
+                         [](Type type) { return llvm::isa<TensorType>(type); });
   });
 
   // Now that the conversion target has been defined, we just need to provide
   // the set of patterns that will lower the Toy operations.
   RewritePatternSet patterns(&getContext());
   patterns.add<AddOpLowering, ConstantOpLowering, FuncOpLowering, MulOpLowering,
                PrintOpLowering, ReturnOpLowering, TransposeOpLowering>(
       &getContext());

mlir/examples/toy/Ch6/mlir/LowerToLLVM.cpp

 class PrintOpLowering : public ConversionPattern {
 public:
   explicit PrintOpLowering(MLIRContext *context)
       : ConversionPattern(toy::PrintOp::getOperationName(), 1, context) {}
 
   LogicalResult
   matchAndRewrite(Operation *op, ArrayRef<Value> operands,
                   ConversionPatternRewriter &rewriter) const override {
-    auto memRefType = (*op->operand_type_begin()).cast<MemRefType>();
+    auto memRefType = llvm::cast<MemRefType>((*op->operand_type_begin()));
     auto memRefShape = memRefType.getShape();
     auto loc = op->getLoc();
 
     ModuleOp parentModule = op->getParentOfType<ModuleOp>();
 
     // Get a symbol reference to the printf function, inserting it if necessary.
     auto printfRef = getOrInsertPrintf(rewriter, parentModule);
     Value formatSpecifierCst = getOrCreateGlobalString(

mlir/examples/toy/Ch6/mlir/ShapeInferencePass.cpp

if (!opWorklist.empty()) {
       signalPassFailure();
     }
   }
 
   /// A utility method that returns if the given operation has all of its
   /// operands inferred.
   static bool allOperandsInferred(Operation *op) {
     return llvm::all_of(op->getOperandTypes(), [](Type operandType) {
-      return operandType.isa<RankedTensorType>();
+      return llvm::isa<RankedTensorType>(operandType);
     });
   }
 
   /// A utility method that returns if the given operation has a dynamically
   /// shaped result.
   static bool returnsDynamicShape(Operation *op) {
     return llvm::any_of(op->getResultTypes(), [](Type resultType) {
-      return !resultType.isa<RankedTensorType>();
+      return !llvm::isa<RankedTensorType>(resultType);
     });
   }
 };
 } // namespace
 
 /// Create a Shape Inference pass.
 std::unique_ptr<mlir::Pass> mlir::toy::createShapeInferencePass() {
   return std::make_unique<ShapeInferencePass>();
 }

mlir/examples/toy/Ch7/mlir/Dialect.cpp

static mlir::ParseResult parseBinaryOp(mlir::OpAsmParser &parser,
   Type type;
   if (parser.parseOperandList(operands, /*requiredOperandCount=*/2) ||
       parser.parseOptionalAttrDict(result.attributes) ||
       parser.parseColonType(type))
     return mlir::failure();
 
   // If the type is a function type, it contains the input and result types of
   // this operation.
-  if (FunctionType funcType = type.dyn_cast<FunctionType>()) {
+  if (FunctionType funcType = llvm::dyn_cast<FunctionType>(type)) {
     if (parser.resolveOperands(operands, funcType.getInputs(), operandsLoc,
                                result.operands))
       return mlir::failure();
     result.addTypes(funcType.getResults());
     return mlir::success();
   }
 
   // Otherwise, the parsed type is the type of both operands and results.
void ConstantOp::print(mlir::OpAsmPrinter &printer) {
   printer.printOptionalAttrDict((*this)->getAttrs(), /*elidedAttrs=*/{"value"});
   printer << getValue();
 }
 
 /// Verify that the given attribute value is valid for the given type.
 static mlir::LogicalResult verifyConstantForType(mlir::Type type,
                                                  mlir::Attribute opaqueValue,
                                                  mlir::Operation *op) {
-  if (type.isa<mlir::TensorType>()) {
+  if (llvm::isa<mlir::TensorType>(type)) {
     // Check that the value is an elements attribute.
-    auto attrValue = opaqueValue.dyn_cast<mlir::DenseFPElementsAttr>();
+    auto attrValue = llvm::dyn_cast<mlir::DenseFPElementsAttr>(opaqueValue);
     if (!attrValue)
       return op->emitError("constant of TensorType must be initialized by "
                            "a DenseFPElementsAttr, got ")
              << opaqueValue;
 
     // If the return type of the constant is not an unranked tensor, the shape
     // must match the shape of the attribute holding the data.
-    auto resultType = type.dyn_cast<mlir::RankedTensorType>();
+    auto resultType = llvm::dyn_cast<mlir::RankedTensorType>(type);
     if (!resultType)
       return success();
 
     // Check that the rank of the attribute type matches the rank of the
     // constant result type.
-    auto attrType = attrValue.getType().cast<mlir::RankedTensorType>();
+    auto attrType = llvm::cast<mlir::RankedTensorType>(attrValue.getType());
     if (attrType.getRank() != resultType.getRank()) {
       return op->emitOpError("return type must match the one of the attached "
                              "value attribute: ")
              << attrType.getRank() << " != " << resultType.getRank();
     }
 
     // Check that each of the dimensions match between the two types.
     for (int dim = 0, dimE = attrType.getRank(); dim < dimE; ++dim) {
       if (attrType.getShape()[dim] != resultType.getShape()[dim]) {
         return op->emitOpError(
                    "return type shape mismatches its attribute at dimension ")
                << dim << ": " << attrType.getShape()[dim]
                << " != " << resultType.getShape()[dim];
       }
     }
     return mlir::success();
   }
-  auto resultType = type.cast<StructType>();
+  auto resultType = llvm::cast<StructType>(type);
   llvm::ArrayRef<mlir::Type> resultElementTypes = resultType.getElementTypes();
 
   // Verify that the initializer is an Array.
-  auto attrValue = opaqueValue.dyn_cast<ArrayAttr>();
+  auto attrValue = llvm::dyn_cast<ArrayAttr>(opaqueValue);
   if (!attrValue || attrValue.getValue().size() != resultElementTypes.size())
     return op->emitError("constant of StructType must be initialized by an "
                          "ArrayAttr with the same number of elements, got ")
            << opaqueValue;
 
   // Check that each of the elements are valid.
   llvm::ArrayRef<mlir::Attribute> attrElementValues = attrValue.getValue();
   for (const auto it : llvm::zip(resultElementTypes, attrElementValues))

 
 /// Returns true if the given set of input and result types are compatible with
 /// this cast operation. This is required by the `CastOpInterface` to verify
 /// this operation and provide other additional utilities.
 bool CastOp::areCastCompatible(TypeRange inputs, TypeRange outputs) {
   if (inputs.size() != 1 || outputs.size() != 1)
     return false;
   // The inputs must be Tensors with the same element type.
-  TensorType input = inputs.front().dyn_cast<TensorType>();
-  TensorType output = outputs.front().dyn_cast<TensorType>();
+  TensorType input = llvm::dyn_cast<TensorType>(inputs.front());
+  TensorType output = llvm::dyn_cast<TensorType>(outputs.front());
   if (!input || !output || input.getElementType() != output.getElementType())
     return false;
   // The shape is required to match if both types are ranked.
   return !input.hasRank() || !output.hasRank() || input == output;
 }
 
 //===----------------------------------------------------------------------===//
 // FuncOp
mlir::LogicalResult ReturnOp::verify() {
   // If the operation does not have an input, we are done.
   if (!hasOperand())
     return mlir::success();
 
   auto inputType = *operand_type_begin();
   auto resultType = results.front();
 
   // Check that the result type of the function matches the operand type.
-  if (inputType == resultType || inputType.isa<mlir::UnrankedTensorType>() ||
-      resultType.isa<mlir::UnrankedTensorType>())
+  if (inputType == resultType || llvm::isa<mlir::UnrankedTensorType>(inputType) ||
+      llvm::isa<mlir::UnrankedTensorType>(resultType))
     return mlir::success();
 
   return emitError() << "type of return operand (" << inputType
                      << ") doesn't match function result type (" << resultType
                      << ")";
 }
 
 //===----------------------------------------------------------------------===//
 // StructAccessOp
 //===----------------------------------------------------------------------===//
 
 void StructAccessOp::build(mlir::OpBuilder &b, mlir::OperationState &state,
                            mlir::Value input, size_t index) {
   // Extract the result type from the input type.
-  StructType structTy = input.getType().cast<StructType>();
+  StructType structTy = llvm::cast<StructType>(input.getType());
   assert(index < structTy.getNumElementTypes());
   mlir::Type resultType = structTy.getElementTypes()[index];
 
   // Call into the auto-generated build method.
   build(b, state, resultType, input, b.getI64IntegerAttr(index));
 }
 
 mlir::LogicalResult StructAccessOp::verify() {
-  StructType structTy = getInput().getType().cast<StructType>();
+  StructType structTy = llvm::cast<StructType>(getInput().getType());
   size_t indexValue = getIndex();
   if (indexValue >= structTy.getNumElementTypes())
     return emitOpError()
            << "index should be within the range of the input struct type";
   mlir::Type resultType = getResult().getType();
   if (resultType != structTy.getElementTypes()[indexValue])
     return emitOpError() << "must have the same result type as the struct "
                             "element referred to by the index";
   return mlir::success();
 }
 
 //===----------------------------------------------------------------------===//
 // TransposeOp
 //===----------------------------------------------------------------------===//
 
 void TransposeOp::build(mlir::OpBuilder &builder, mlir::OperationState &state,
                         mlir::Value value) {
   state.addTypes(UnrankedTensorType::get(builder.getF64Type()));
   state.addOperands(value);
 }
 
 void TransposeOp::inferShapes() {
-  auto arrayTy = getOperand().getType().cast<RankedTensorType>();
+  auto arrayTy = llvm::cast<RankedTensorType>(getOperand().getType());
   SmallVector<int64_t, 2> dims(llvm::reverse(arrayTy.getShape()));
   getResult().setType(RankedTensorType::get(dims, arrayTy.getElementType()));
 }
 
 mlir::LogicalResult TransposeOp::verify() {
-  auto inputType = getOperand().getType().dyn_cast<RankedTensorType>();
-  auto resultType = getType().dyn_cast<RankedTensorType>();
+  auto inputType = llvm::dyn_cast<RankedTensorType>(getOperand().getType());
+  auto resultType = llvm::dyn_cast<RankedTensorType>(getType());
   if (!inputType || !resultType)
     return mlir::success();
 
   auto inputShape = inputType.getShape();
   if (!std::equal(inputShape.begin(), inputShape.end(),
                   resultType.getShape().rbegin())) {
     return emitError()
            << "expected result shape to be a transpose of the input";
mlir::Type ToyDialect::parseType(mlir::DialectAsmParser &parser) const {
   do {
     // Parse the current element type.
     SMLoc typeLoc = parser.getCurrentLocation();
     mlir::Type elementType;
     if (parser.parseType(elementType))
       return nullptr;
 
     // Check that the type is either a TensorType or another StructType.
-    if (!elementType.isa<mlir::TensorType, StructType>()) {
+    if (!llvm::isa<mlir::TensorType, StructType>(elementType)) {
       parser.emitError(typeLoc, "element type for a struct must either "
                                 "be a TensorType or a StructType, got: ")
           << elementType;
       return Type();
     }
     elementTypes.push_back(elementType);
 
     // Parse the optional: `,`
   } while (succeeded(parser.parseOptionalComma()));
 
   // Parse: `>`
   if (parser.parseGreater())
     return Type();
   return StructType::get(elementTypes);
 }
 
 /// Print an instance of a type registered to the toy dialect.
 void ToyDialect::printType(mlir::Type type,
                            mlir::DialectAsmPrinter &printer) const {
   // Currently the only toy type is a struct type.
-  StructType structType = type.cast<StructType>();
+  StructType structType = llvm::cast<StructType>(type);
 
   // Print the struct type according to the parser format.
   printer << "struct<";
   llvm::interleaveComma(structType.getElementTypes(), printer);
   printer << '>';
 }
 
 //===----------------------------------------------------------------------===//
#include "toy/Ops.cpp.inc"
   addInterfaces<ToyInlinerInterface>();
   addTypes<StructType>();
 }
 
 mlir::Operation *ToyDialect::materializeConstant(mlir::OpBuilder &builder,
                                                  mlir::Attribute value,
                                                  mlir::Type type,
                                                  mlir::Location loc) {
-  if (type.isa<StructType>())
+  if (llvm::isa<StructType>(type))
     return builder.create<StructConstantOp>(loc, type,
-                                            value.cast<mlir::ArrayAttr>());
+                                            llvm::cast<mlir::ArrayAttr>(value));
   return builder.create<ConstantOp>(loc, type,
-                                    value.cast<mlir::DenseElementsAttr>());
+                                    llvm::cast<mlir::DenseElementsAttr>(value));
 }

mlir/examples/toy/Ch7/mlir/LowerToAffineLoops.cpp

 /// induction variables for the iteration. It returns a value to store at the
 /// current index of the iteration.
 using LoopIterationFn = function_ref<Value(
     OpBuilder &rewriter, ValueRange memRefOperands, ValueRange loopIvs)>;
 
 static void lowerOpToLoops(Operation *op, ValueRange operands,
                            PatternRewriter &rewriter,
                            LoopIterationFn processIteration) {
-  auto tensorType = (*op->result_type_begin()).cast<RankedTensorType>();
+  auto tensorType = llvm::cast<RankedTensorType>((*op->result_type_begin()));
   auto loc = op->getLoc();
 
   // Insert an allocation and deallocation for the result of this operation.
   auto memRefType = convertTensorToMemRef(tensorType);
   auto alloc = insertAllocAndDealloc(memRefType, loc, rewriter);
 
   // Create a nest of affine loops, with one loop per dimension of the shape.
   // The buildAffineLoopNest function takes a callback that is used to construct
struct ConstantOpLowering : public OpRewritePattern<toy::ConstantOp> {
 
   LogicalResult matchAndRewrite(toy::ConstantOp op,
                                 PatternRewriter &rewriter) const final {
     DenseElementsAttr constantValue = op.getValue();
     Location loc = op.getLoc();
 
     // When lowering the constant operation, we allocate and assign the constant
     // values to a corresponding memref allocation.
-    auto tensorType = op.getType().cast<RankedTensorType>();
+    auto tensorType = llvm::cast<RankedTensorType>(op.getType());
     auto memRefType = convertTensorToMemRef(tensorType);
     auto alloc = insertAllocAndDealloc(memRefType, loc, rewriter);
 
     // We will be generating constant indices up-to the largest dimension.
     // Create these constants up-front to avoid large amounts of redundant
     // operations.
     auto valueShape = memRefType.getShape();
     SmallVector<Value, 8> constantIndices;
void ToyToAffineLoweringPass::runOnOperation() {
   // if any of these operations are *not* converted. Given that we actually want
   // a partial lowering, we explicitly mark the Toy operations that don't want
   // to lower, `toy.print`, as `legal`. `toy.print` will still need its operands
   // to be updated though (as we convert from TensorType to MemRefType), so we
   // only treat it as `legal` if its operands are legal.
   target.addIllegalDialect<toy::ToyDialect>();
   target.addDynamicallyLegalOp<toy::PrintOp>([](toy::PrintOp op) {
     return llvm::none_of(op->getOperandTypes(),
-                         [](Type type) { return type.isa<TensorType>(); });
+                         [](Type type) { return llvm::isa<TensorType>(type); });
   });
 
   // Now that the conversion target has been defined, we just need to provide
   // the set of patterns that will lower the Toy operations.
   RewritePatternSet patterns(&getContext());
   patterns.add<AddOpLowering, ConstantOpLowering, FuncOpLowering, MulOpLowering,
                PrintOpLowering, ReturnOpLowering, TransposeOpLowering>(
       &getContext());

mlir/examples/toy/Ch7/mlir/LowerToLLVM.cpp

 class PrintOpLowering : public ConversionPattern {
 public:
   explicit PrintOpLowering(MLIRContext *context)
       : ConversionPattern(toy::PrintOp::getOperationName(), 1, context) {}
 
   LogicalResult
   matchAndRewrite(Operation *op, ArrayRef<Value> operands,
                   ConversionPatternRewriter &rewriter) const override {
-    auto memRefType = (*op->operand_type_begin()).cast<MemRefType>();
+    auto memRefType = llvm::cast<MemRefType>((*op->operand_type_begin()));
     auto memRefShape = memRefType.getShape();
     auto loc = op->getLoc();
 
     ModuleOp parentModule = op->getParentOfType<ModuleOp>();
 
     // Get a symbol reference to the printf function, inserting it if necessary.
     auto printfRef = getOrInsertPrintf(rewriter, parentModule);
     Value formatSpecifierCst = getOrCreateGlobalString(

mlir/examples/toy/Ch7/mlir/ShapeInferencePass.cpp

if (!opWorklist.empty()) {
       signalPassFailure();
     }
   }
 
   /// A utility method that returns if the given operation has all of its
   /// operands inferred.
   static bool allOperandsInferred(Operation *op) {
     return llvm::all_of(op->getOperandTypes(), [](Type operandType) {
-      return operandType.isa<RankedTensorType>();
+      return llvm::isa<RankedTensorType>(operandType);
     });
   }
 
   /// A utility method that returns if the given operation has a dynamically
   /// shaped result.
   static bool returnsDynamicShape(Operation *op) {
     return llvm::any_of(op->getResultTypes(), [](Type resultType) {
-      return !resultType.isa<RankedTensorType>();
+      return !llvm::isa<RankedTensorType>(resultType);
     });
   }
 };
 } // namespace
 
 /// Create a Shape Inference pass.
 std::unique_ptr<mlir::Pass> mlir::toy::createShapeInferencePass() {
   return std::make_unique<ShapeInferencePass>();
 }

mlir/examples/toy/Ch7/mlir/ToyCombine.cpp

 /// Fold constants.
 OpFoldResult ConstantOp::fold(FoldAdaptor adaptor) { return getValue(); }
 
 /// Fold struct constants.
 OpFoldResult StructConstantOp::fold(FoldAdaptor adaptor) { return getValue(); }
 
 /// Fold simple struct access operations that access into a constant.
 OpFoldResult StructAccessOp::fold(FoldAdaptor adaptor) {
-  auto structAttr = adaptor.getInput().dyn_cast_or_null<mlir::ArrayAttr>();
+  auto structAttr =
+      llvm::dyn_cast_if_present<mlir::ArrayAttr>(adaptor.getInput());
   if (!structAttr)
     return nullptr;
 
   size_t elementIndex = getIndex();
   return structAttr[elementIndex];
 }
 
 /// This is an example of a c++ rewrite pattern for the TransposeOp. It

mlir/include/mlir/Debug/BreakpointManagers/FileLineColLocBreakpointManager.h

 /// FileLineColLocBreakpoint. It'll extract the location from the action context
 /// looking for a FileLineColLocation, and match it against the registered
 /// breakpoints.
 class FileLineColLocBreakpointManager
     : public BreakpointManagerBase<FileLineColLocBreakpointManager> {
 public:
   Breakpoint *match(const Action &action) const override {
     for (const IRUnit &unit : action.getContextIRUnits()) {
-      if (auto *op = unit.dyn_cast<Operation *>()) {
+      if (auto *op = llvm::dyn_cast_if_present<Operation *>(unit)) {
         if (auto match = matchFromLocation(op->getLoc()))
           return *match;
         continue;
       }
-      if (auto *block = unit.dyn_cast<Block *>()) {
+      if (auto *block = llvm::dyn_cast_if_present<Block *>(unit)) {
         for (auto &op : block->getOperations()) {
           if (auto match = matchFromLocation(op.getLoc()))
             return *match;
         }
         continue;
       }
-      if (Region *region = unit.dyn_cast<Region *>()) {
+      if (Region *region = llvm::dyn_cast_if_present<Region *>(unit)) {
         if (auto match = matchFromLocation(region->getLoc()))
           return *match;
         continue;
       }
     }
     return {};
   }
 

mlir/include/mlir/Dialect/GPU/IR/GPUBase.td

 def GPU_MMAMemRef : MemRefOf<[I8, I32, F16, F32, VectorOfRankAndType<[1], [I8, I32, F16, F32]>]>;
 
 class MMAMatrixOf<list<Type> allowedTypes> :
   ContainerType<AnyTypeOf<allowedTypes>, IsMMAMatrixTypePred,
   "::llvm::cast<::mlir::gpu::MMAMatrixType>($_self).getElementType()",
   "gpu.mma_matrix", "::mlir::gpu::MMAMatrixType">;
 
 // Types for all sparse handles.
-def GPU_SparseEnvHandle : 
-  DialectType<GPU_Dialect, 
-    CPred<"$_self.isa<::mlir::gpu::SparseEnvHandleType>()">, 
-    "sparse environment handle type">, 
+def GPU_SparseEnvHandle :
+  DialectType<GPU_Dialect,
+    CPred<"llvm::isa<::mlir::gpu::SparseEnvHandleType>($_self)">,
+    "sparse environment handle type">,
   BuildableType<"mlir::gpu::SparseEnvHandleType::get($_builder.getContext())">;
 
-def GPU_SparseDnVecHandle : 
-  DialectType<GPU_Dialect, 
-    CPred<"$_self.isa<::mlir::gpu::SparseDnVecHandleType>()">, 
+def GPU_SparseDnVecHandle :
+  DialectType<GPU_Dialect,
+    CPred<"llvm::isa<::mlir::gpu::SparseDnVecHandleType>($_self)">,
     "dense vector handle type">,
   BuildableType<"mlir::gpu::SparseDnVecHandleType::get($_builder.getContext())">;
 
-def GPU_SparseDnMatHandle : 
-  DialectType<GPU_Dialect, 
-    CPred<"$_self.isa<::mlir::gpu::SparseDnMatHandleType>()">, 
+def GPU_SparseDnMatHandle :
+  DialectType<GPU_Dialect,
+    CPred<"llvm::isa<::mlir::gpu::SparseDnMatHandleType>($_self)">,
     "dense matrix handle type">,
   BuildableType<"mlir::gpu::SparseDnMatHandleType::get($_builder.getContext())">;
 
-def GPU_SparseSpMatHandle : 
-  DialectType<GPU_Dialect, 
-    CPred<"$_self.isa<::mlir::gpu::SparseSpMatHandleType>()">, 
+def GPU_SparseSpMatHandle :
+  DialectType<GPU_Dialect,
+    CPred<"llvm::isa<::mlir::gpu::SparseSpMatHandleType>($_self)">,
     "sparse matrix handle type">,
   BuildableType<"mlir::gpu::SparseSpMatHandleType::get($_builder.getContext())">;
 
 //===----------------------------------------------------------------------===//
 // GPU Interfaces.
 //===----------------------------------------------------------------------===//
 
 def GPU_AsyncOpInterface : OpInterface<"AsyncOpInterface"> {

mlir/include/mlir/Dialect/OpenMP/OpenMPOpsInterfaces.td

InterfaceMethod<
       /*description=*/[{
         Returns the DeclareTargetDeviceType segment of the DeclareTarget attribute if it
         exists on the current operation. Otherwise it returns null.
       }],
       /*retTy=*/"mlir::omp::DeclareTargetDeviceType",
       /*methodName=*/"getDeclareTargetDeviceType",
       (ins), [{}], [{
         if (mlir::Attribute dTar = $_op->getAttr("omp.declare_target"))
-          if (auto dAttr = dTar.dyn_cast_or_null<mlir::omp::DeclareTargetAttr>())
+          if (auto dAttr = llvm::dyn_cast_or_null<mlir::omp::DeclareTargetAttr>(dTar))
             return dAttr.getDeviceType().getValue();
         return {};
       }]>,
       InterfaceMethod<
       /*description=*/[{
         Returns the DeclareTargetCaptureClause segment of the DeclareTarget attribute if it
         exists on the current operation. Otherwise it returns null.
       }],
       /*retTy=*/"mlir::omp::DeclareTargetCaptureClause",
       /*methodName=*/"getDeclareTargetCaptureClause",
       (ins), [{}], [{
         if (mlir::Attribute dTar = $_op->getAttr("omp.declare_target"))
-          if (auto dAttr = dTar.dyn_cast_or_null<mlir::omp::DeclareTargetAttr>())
+          if (auto dAttr = llvm::dyn_cast_or_null<mlir::omp::DeclareTargetAttr>(dTar))
             return dAttr.getCaptureClause().getValue();
         return {};
       }]>
   ];
 }
 
 def OffloadModuleInterface : OpInterface<"OffloadModuleInterface"> {
   let description = [{

mlir/include/mlir/IR/BuiltinTypes.h

public:
 
   /// Return true if the specified element type is ok in a tensor.
   static bool isValidElementType(Type type);
 
   /// Methods for support type inquiry through isa, cast, and dyn_cast.
   static bool classof(Type type);
 
   /// Allow implicit conversion to ShapedType.
-  operator ShapedType() const { return cast<ShapedType>(); }
+  operator ShapedType() const { return llvm::cast<ShapedType>(*this); }
 };
 
 //===----------------------------------------------------------------------===//
 // BaseMemRefType
 //===----------------------------------------------------------------------===//
 
 /// This class provides a shared interface for ranked and unranked memref types.
 /// Note: This class attaches the ShapedType trait to act as a mixin to
public:
   /// Returns the memory space in which data referred to by this memref resides.
   Attribute getMemorySpace() const;
 
   /// [deprecated] Returns the memory space in old raw integer representation.
   /// New `Attribute getMemorySpace()` method should be used instead.
   unsigned getMemorySpaceAsInt() const;
 
   /// Allow implicit conversion to ShapedType.
-  operator ShapedType() const { return cast<ShapedType>(); }
+  operator ShapedType() const { return llvm::cast<ShapedType>(*this); }
 };
 
 } // namespace mlir
 
 //===----------------------------------------------------------------------===//
 // Tablegen Type Declarations
 //===----------------------------------------------------------------------===//
 

mlir/include/mlir/IR/TypeRange.h

static const mlir::Type *getEmptyKeyPointer() {
     return DenseMapInfo<mlir::Type *>::getEmptyKey();
   }
 
   static const mlir::Type *getTombstoneKeyPointer() {
     return DenseMapInfo<mlir::Type *>::getTombstoneKey();
   }
 
   static bool isEmptyKey(mlir::TypeRange range) {
-    if (const auto *type = range.getBase().dyn_cast<const mlir::Type *>())
+    if (const auto *type =
+            llvm::dyn_cast_if_present<const mlir::Type *>(range.getBase()))
       return type == getEmptyKeyPointer();
     return false;
   }
 
   static bool isTombstoneKey(mlir::TypeRange range) {
-    if (const auto *type = range.getBase().dyn_cast<const mlir::Type *>())
+    if (const auto *type =
+            llvm::dyn_cast_if_present<const mlir::Type *>(range.getBase()))
       return type == getTombstoneKeyPointer();
     return false;
   }
 };
 
 } // namespace llvm
 
 #endif // MLIR_IR_TYPERANGE_H

mlir/include/mlir/Interfaces/SideEffectInterfaces.h

EffectInstance(EffectT *effect, SymbolRefAttr symbol, Attribute parameters,
       : effect(effect), resource(resource), value(symbol),
         parameters(parameters) {}
 
   /// Return the effect being applied.
   EffectT *getEffect() const { return effect; }
 
   /// Return the value the effect is applied on, or nullptr if there isn't a
   /// known value being affected.
-  Value getValue() const { return value ? value.dyn_cast<Value>() : Value(); }
+  Value getValue() const { return value ? llvm::dyn_cast_if_present<Value>(value) : Value(); }
 
   /// Return the symbol reference the effect is applied on, or nullptr if there
   /// isn't a known smbol being affected.
   SymbolRefAttr getSymbolRef() const {
-    return value ? value.dyn_cast<SymbolRefAttr>() : SymbolRefAttr();
+    return value ? llvm::dyn_cast_if_present<SymbolRefAttr>(value) : SymbolRefAttr();
   }
 
   /// Return the resource that the effect applies to.
   Resource *getResource() const { return resource; }
 
   /// Return the parameters of the effect, if any.
   Attribute getParameters() const { return parameters; }
 

mlir/include/mlir/Pass/AnalysisManager.h

struct NestedAnalysisMap {
   Operation *getOperation() const { return analyses.getOperation(); }
 
   /// Invalidate any non preserved analyses.
   void invalidate(const PreservedAnalyses &pa);
 
   /// Returns the parent analysis map for this analysis map, or null if this is
   /// the top-level map.
   const NestedAnalysisMap *getParent() const {
-    return parentOrInstrumentor.dyn_cast<NestedAnalysisMap *>();
+    return llvm::dyn_cast_if_present<NestedAnalysisMap *>(parentOrInstrumentor);
   }
 
   /// Returns a pass instrumentation object for the current operation. This
   /// value may be null.
   PassInstrumentor *getPassInstrumentor() const {
     if (auto *parent = getParent())
       return parent->getPassInstrumentor();
     return parentOrInstrumentor.get<PassInstrumentor *>();

mlir/lib/Analysis/DataFlow/ConstantPropagationAnalysis.cpp

void SparseConstantPropagation::visitOperation(
 
   // Merge the fold results into the lattice for this operation.
   assert(foldResults.size() == op->getNumResults() && "invalid result size");
   for (const auto it : llvm::zip(results, foldResults)) {
     Lattice<ConstantValue> *lattice = std::get<0>(it);
 
     // Merge in the result of the fold, either a constant or a value.
     OpFoldResult foldResult = std::get<1>(it);
-    if (Attribute attr = foldResult.dyn_cast<Attribute>()) {
+    if (Attribute attr = llvm::dyn_cast_if_present<Attribute>(foldResult)) {
       LLVM_DEBUG(llvm::dbgs() << "Folded to constant: " << attr << "\n");
       propagateIfChanged(lattice,
                          lattice->join(ConstantValue(attr, op->getDialect())));
     } else {
       LLVM_DEBUG(llvm::dbgs()
                  << "Folded to value: " << foldResult.get<Value>() << "\n");
       AbstractSparseDataFlowAnalysis::join(
           lattice, *getLatticeElement(foldResult.get<Value>()));

mlir/lib/Analysis/DataFlow/DeadCodeAnalysis.cpp

ChangeResult Executable::setToLive() {
   return ChangeResult::Change;
 }
 
 void Executable::print(raw_ostream &os) const {
   os << (live ? "live" : "dead");
 }
 
 void Executable::onUpdate(DataFlowSolver *solver) const {
-  if (auto *block = point.dyn_cast<Block *>()) {
+  if (auto *block = llvm::dyn_cast_if_present<Block *>(point)) {
     // Re-invoke the analyses on the block itself.
     for (DataFlowAnalysis *analysis : subscribers)
       solver->enqueue({block, analysis});
     // Re-invoke the analyses on all operations in the block.
     for (DataFlowAnalysis *analysis : subscribers)
       for (Operation &op : *block)
         solver->enqueue({&op, analysis});
-  } else if (auto *programPoint = point.dyn_cast<GenericProgramPoint *>()) {
+  } else if (auto *programPoint = llvm::dyn_cast_if_present<GenericProgramPoint *>(point)) {
     // Re-invoke the analysis on the successor block.
     if (auto *edge = dyn_cast<CFGEdge>(programPoint)) {
       for (DataFlowAnalysis *analysis : subscribers)
         solver->enqueue({edge->getTo(), analysis});
     }
   }
 }
 
for (Region &region : op->getRegions()) {
     auto *state = getOrCreate<Executable>(&region.front());
     propagateIfChanged(state, state->setToLive());
   }
 }
 
 LogicalResult DeadCodeAnalysis::visit(ProgramPoint point) {
   if (point.is<Block *>())
     return success();
-  auto *op = point.dyn_cast<Operation *>();
+  auto *op = llvm::dyn_cast_if_present<Operation *>(point);
   if (!op)
     return emitError(point.getLoc(), "unknown program point kind");
 
   // If the parent block is not executable, there is nothing to do.
   if (!getOrCreate<Executable>(op->getBlock())->isLive())
     return success();
 
   // We have a live call op. Add this as a live predecessor of the callee.

mlir/lib/Analysis/DataFlow/DenseAnalysis.cpp

for (Block &block : region) {
         if (failed(initialize(&op)))
           return failure();
     }
   }
   return success();
 }
 
 LogicalResult AbstractDenseDataFlowAnalysis::visit(ProgramPoint point) {
-  if (auto *op = point.dyn_cast<Operation *>())
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(point))
     processOperation(op);
-  else if (auto *block = point.dyn_cast<Block *>())
+  else if (auto *block = llvm::dyn_cast_if_present<Block *>(point))
     visitBlock(block);
   else
     return failure();
   return success();
 }
 
 void AbstractDenseDataFlowAnalysis::processOperation(Operation *op) {
   // If the containing block is not executable, bail out.

mlir/lib/Analysis/DataFlow/IntegerRangeAnalysis.cpp

void IntegerRangeAnalysis::visitNonControlFlowArguments(
   auto getLoopBoundFromFold = [&](std::optional<OpFoldResult> loopBound,
                                   Type boundType, bool getUpper) {
     unsigned int width = ConstantIntRanges::getStorageBitwidth(boundType);
     if (loopBound.has_value()) {
       if (loopBound->is<Attribute>()) {
         if (auto bound =
                 dyn_cast_or_null<IntegerAttr>(loopBound->get<Attribute>()))
           return bound.getValue();
-      } else if (auto value = loopBound->dyn_cast<Value>()) {
+      } else if (auto value = llvm::dyn_cast_if_present<Value>(*loopBound)) {
         const IntegerValueRangeLattice *lattice =
             getLatticeElementFor(op, value);
         if (lattice != nullptr)
           return getUpper ? lattice->getValue().getValue().smax()
                           : lattice->getValue().getValue().smin();
       }
     }
     // Given the results of getConstant{Lower,Upper}Bound()

mlir/lib/Analysis/DataFlow/SparseAnalysis.cpp

for (Block &block : region) {
           return failure();
     }
   }
 
   return success();
 }
 
 LogicalResult AbstractSparseDataFlowAnalysis::visit(ProgramPoint point) {
-  if (Operation *op = point.dyn_cast<Operation *>())
+  if (Operation *op = llvm::dyn_cast_if_present<Operation *>(point))
     visitOperation(op);
-  else if (Block *block = point.dyn_cast<Block *>())
+  else if (Block *block = llvm::dyn_cast_if_present<Block *>(point))
     visitBlock(block);
   else
     return failure();
   return success();
 }
 
 void AbstractSparseDataFlowAnalysis::visitOperation(Operation *op) {
   // Exit early on operations with no results.
for (Operation *op : predecessors->getKnownPredecessors()) {
     }
 
     ValueRange inputs = predecessors->getSuccessorInputs(op);
     assert(inputs.size() == operands->size() &&
            "expected the same number of successor inputs as operands");
 
     unsigned firstIndex = 0;
     if (inputs.size() != lattices.size()) {
-      if (point.dyn_cast<Operation *>()) {
+      if (llvm::dyn_cast_if_present<Operation *>(point)) {
         if (!inputs.empty())
           firstIndex = cast<OpResult>(inputs.front()).getResultNumber();
         visitNonControlFlowArgumentsImpl(
             branch,
             RegionSuccessor(
                 branch->getResults().slice(firstIndex, inputs.size())),
             lattices, firstIndex);
       } else {
for (Block &block : region) {
           return failure();
     }
   }
   return success();
 }
 
 LogicalResult
 AbstractSparseBackwardDataFlowAnalysis::visit(ProgramPoint point) {
-  if (Operation *op = point.dyn_cast<Operation *>())
+  if (Operation *op = llvm::dyn_cast_if_present<Operation *>(point))
     visitOperation(op);
-  else if (point.dyn_cast<Block *>())
+  else if (llvm::dyn_cast_if_present<Block *>(point))
     // For backward dataflow, we don't have to do any work for the blocks
     // themselves. CFG edges between blocks are processed by the BranchOp
     // logic in `visitOperation`, and entry blocks for functions are tied
     // to the CallOp arguments by visitOperation.
     return success();
   else
     return failure();
   return success();

mlir/lib/Analysis/DataFlowFramework.cpp

 // ProgramPoint
 //===----------------------------------------------------------------------===//
 
 void ProgramPoint::print(raw_ostream &os) const {
   if (isNull()) {
     os << "<NULL POINT>";
     return;
   }
-  if (auto *programPoint = dyn_cast<GenericProgramPoint *>())
+  if (auto *programPoint = llvm::dyn_cast<GenericProgramPoint *>(*this))
     return programPoint->print(os);
-  if (auto *op = dyn_cast<Operation *>())
+  if (auto *op = llvm::dyn_cast<Operation *>(*this))
     return op->print(os);
-  if (auto value = dyn_cast<Value>())
+  if (auto value = llvm::dyn_cast<Value>(*this))
     return value.print(os);
   return get<Block *>()->print(os);
 }
 
 Location ProgramPoint::getLoc() const {
-  if (auto *programPoint = dyn_cast<GenericProgramPoint *>())
+  if (auto *programPoint = llvm::dyn_cast<GenericProgramPoint *>(*this))
     return programPoint->getLoc();
-  if (auto *op = dyn_cast<Operation *>())
+  if (auto *op = llvm::dyn_cast<Operation *>(*this))
     return op->getLoc();
-  if (auto value = dyn_cast<Value>())
+  if (auto value = llvm::dyn_cast<Value>(*this))
     return value.getLoc();
   return get<Block *>()->getParent()->getLoc();
 }
 
 //===----------------------------------------------------------------------===//
 // DataFlowSolver
 //===----------------------------------------------------------------------===//
 

mlir/lib/AsmParser/Parser.cpp

if (parseLocationAlias(directLoc))
       return failure();
   } else if (parseLocationInstance(directLoc)) {
     return failure();
   }
 
   if (parseToken(Token::r_paren, "expected ')' in location"))
     return failure();
 
-  if (auto *op = opOrArgument.dyn_cast<Operation *>())
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(opOrArgument))
     op->setLoc(directLoc);
   else
     opOrArgument.get<BlockArgument>().setLoc(directLoc);
   return success();
 }
 
 //===----------------------------------------------------------------------===//
 // Region Parsing

mlir/lib/CAPI/Interfaces/Interfaces.cpp

SmallVector<Value> unwrapOperands(intptr_t nOperands, MlirValue *operands) {
   SmallVector<Value> unwrappedOperands;
   (void)unwrapList(nOperands, operands, unwrappedOperands);
   return unwrappedOperands;
 }
 
 DictionaryAttr unwrapAttributes(MlirAttribute attributes) {
   DictionaryAttr attributeDict;
   if (!mlirAttributeIsNull(attributes))
-    attributeDict = unwrap(attributes).cast<DictionaryAttr>();
+    attributeDict = llvm::cast<DictionaryAttr>(unwrap(attributes));
   return attributeDict;
 }
 
 SmallVector<std::unique_ptr<Region>> unwrapRegions(intptr_t nRegions,
                                                    MlirRegion *regions) {
   // Create a vector of unique pointers to regions and make sure they are not
   // deleted when exiting the scope. This is a hack caused by C++ API expecting
   // an list of unique pointers to regions (without ownership transfer

mlir/lib/Conversion/GPUCommon/GPUToLLVMConversion.cpp

LogicalResult ConvertSetDefaultDeviceOpToGpuRuntimeCallPattern::matchAndRewrite(
   rewriter.replaceOp(op, {});
   return success();
 }
 
 // Returns the element type of the defining spmat op.
 // TODO: safer and more flexible to store data type in actual op instead?
 static Type getSpMatElemType(Value spMat) {
   if (auto op = spMat.getDefiningOp<gpu::CreateCooOp>())
-    return op.getValues().getType().cast<MemRefType>().getElementType();
+    return llvm::cast<MemRefType>(op.getValues().getType()).getElementType();
   if (auto op = spMat.getDefiningOp<gpu::CreateCsrOp>())
-    return op.getValues().getType().cast<MemRefType>().getElementType();
+    return llvm::cast<MemRefType>(op.getValues().getType()).getElementType();
   llvm_unreachable("cannot find spmat def");
 }
 
 LogicalResult ConvertCreateSparseEnvOpToGpuRuntimeCallPattern::matchAndRewrite(
     gpu::CreateSparseEnvOp op, OpAdaptor adaptor,
     ConversionPatternRewriter &rewriter) const {
   if (failed(areAllLLVMTypes(op, adaptor.getOperands(), rewriter)) ||
       failed(isAsyncWithOneDependency(rewriter, op)))
if (failed(areAllLLVMTypes(op, adaptor.getOperands(), rewriter)) ||
       failed(isAsyncWithOneDependency(rewriter, op)))
     return failure();
   Location loc = op.getLoc();
   auto stream = adaptor.getAsyncDependencies().front();
   Value pVec =
       MemRefDescriptor(adaptor.getMemref()).allocatedPtr(rewriter, loc);
   if (!getTypeConverter()->useOpaquePointers())
     pVec = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pVec);
-  Type dType = op.getMemref().getType().cast<MemRefType>().getElementType();
+  Type dType = llvm::cast<MemRefType>(op.getMemref().getType()).getElementType();
   auto dw = rewriter.create<LLVM::ConstantOp>(loc, llvmInt32Type,
                                               dType.getIntOrFloatBitWidth());
   auto handle =
       createDnVecCallBuilder
           .create(loc, rewriter, {adaptor.getSize(), pVec, dw, stream})
           .getResult();
   rewriter.replaceOp(op, {handle, stream});
   return success();
if (failed(areAllLLVMTypes(op, adaptor.getOperands(), rewriter)) ||
       failed(isAsyncWithOneDependency(rewriter, op)))
     return failure();
   Location loc = op.getLoc();
   auto stream = adaptor.getAsyncDependencies().front();
   Value pMat =
       MemRefDescriptor(adaptor.getMemref()).allocatedPtr(rewriter, loc);
   if (!getTypeConverter()->useOpaquePointers())
     pMat = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pMat);
-  Type dType = op.getMemref().getType().cast<MemRefType>().getElementType();
+  Type dType = llvm::cast<MemRefType>(op.getMemref().getType()).getElementType();
   auto dw = rewriter.create<LLVM::ConstantOp>(loc, llvmInt32Type,
                                               dType.getIntOrFloatBitWidth());
   auto handle =
       createDnMatCallBuilder
           .create(loc, rewriter,
                   {adaptor.getRows(), adaptor.getCols(), pMat, dw, stream})
           .getResult();
   rewriter.replaceOp(op, {handle, stream});
Value pColIdxs =
       MemRefDescriptor(adaptor.getColIdxs()).allocatedPtr(rewriter, loc);
   Value pValues =
       MemRefDescriptor(adaptor.getValues()).allocatedPtr(rewriter, loc);
   if (!getTypeConverter()->useOpaquePointers()) {
     pRowIdxs = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pRowIdxs);
     pColIdxs = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pColIdxs);
     pValues = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pValues);
   }
-  Type iType = op.getColIdxs().getType().cast<MemRefType>().getElementType();
-  Type dType = op.getValues().getType().cast<MemRefType>().getElementType();
+  Type iType = llvm::cast<MemRefType>(op.getColIdxs().getType()).getElementType();
+  Type dType = llvm::cast<MemRefType>(op.getValues().getType()).getElementType();
   auto iw = rewriter.create<LLVM::ConstantOp>(
       loc, llvmInt32Type, iType.isIndex() ? 64 : iType.getIntOrFloatBitWidth());
   auto dw = rewriter.create<LLVM::ConstantOp>(loc, llvmInt32Type,
                                               dType.getIntOrFloatBitWidth());
   auto handle =
       createCooCallBuilder
           .create(loc, rewriter,
                   {adaptor.getRows(), adaptor.getCols(), adaptor.getNnz(),
Value pColIdxs =
       MemRefDescriptor(adaptor.getColIdxs()).allocatedPtr(rewriter, loc);
   Value pValues =
       MemRefDescriptor(adaptor.getValues()).allocatedPtr(rewriter, loc);
   if (!getTypeConverter()->useOpaquePointers()) {
     pRowPos = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pRowPos);
     pColIdxs = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pColIdxs);
     pValues = rewriter.create<LLVM::BitcastOp>(loc, llvmPointerType, pValues);
   }
-  Type pType = op.getRowPos().getType().cast<MemRefType>().getElementType();
-  Type iType = op.getColIdxs().getType().cast<MemRefType>().getElementType();
-  Type dType = op.getValues().getType().cast<MemRefType>().getElementType();
+  Type pType = llvm::cast<MemRefType>(op.getRowPos().getType()).getElementType();
+  Type iType = llvm::cast<MemRefType>(op.getColIdxs().getType()).getElementType();
+  Type dType = llvm::cast<MemRefType>(op.getValues().getType()).getElementType();
   auto pw = rewriter.create<LLVM::ConstantOp>(
       loc, llvmInt32Type, pType.isIndex() ? 64 : pType.getIntOrFloatBitWidth());
   auto iw = rewriter.create<LLVM::ConstantOp>(
       loc, llvmInt32Type, iType.isIndex() ? 64 : iType.getIntOrFloatBitWidth());
   auto dw = rewriter.create<LLVM::ConstantOp>(loc, llvmInt32Type,
                                               dType.getIntOrFloatBitWidth());
   auto handle =
       createCsrCallBuilder

mlir/lib/Conversion/LLVMCommon/TypeConverter.cpp

LLVMTypeConverter::getMemRefAddressSpace(BaseMemRefType type) {
   if (!type.getMemorySpace()) // Default memory space -> 0.
     return 0;
   std::optional<Attribute> converted =
       convertTypeAttribute(type, type.getMemorySpace());
   if (!converted)
     return failure();
   if (!(*converted)) // Conversion to default is 0.
     return 0;
-  if (auto explicitSpace = converted->dyn_cast_or_null<IntegerAttr>())
+  if (auto explicitSpace = llvm::dyn_cast_if_present<IntegerAttr>(*converted))
     return explicitSpace.getInt();
   return failure();
 }
 
 // Check if a memref type can be converted to a bare pointer.
 bool LLVMTypeConverter::canConvertToBarePtr(BaseMemRefType type) {
   if (isa<UnrankedMemRefType>(type))
     // Unranked memref is not supported in the bare pointer calling convention.

mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp

matchAndRewrite(memref::GlobalOp global, OpAdaptor adaptor,
 
     Type arrayTy = convertGlobalMemrefTypeToLLVM(type, *getTypeConverter());
 
     LLVM::Linkage linkage =
         global.isPublic() ? LLVM::Linkage::External : LLVM::Linkage::Private;
 
     Attribute initialValue = nullptr;
     if (!global.isExternal() && !global.isUninitialized()) {
-      auto elementsAttr = global.getInitialValue()->cast<ElementsAttr>();
+      auto elementsAttr = llvm::cast<ElementsAttr>(*global.getInitialValue());
       initialValue = elementsAttr;
 
       // For scalar memrefs, the global variable created is of the element type,
       // so unpack the elements attribute to extract the value.
       if (type.getRank() == 0)
         initialValue = elementsAttr.getSplatValue<Attribute>();
     }
 

mlir/lib/Conversion/PDLToPDLInterp/PDLToPDLInterp.cpp

case Predicates::OperationNameQuestion: {
     builder.create<pdl_interp::CheckOperationNameOp>(
         loc, val, opNameAnswer->getValue().getStringRef(), success, failure);
     break;
   }
   case Predicates::TypeQuestion: {
     auto *ans = cast<TypeAnswer>(answer);
     if (isa<pdl::RangeType>(val.getType()))
       builder.create<pdl_interp::CheckTypesOp>(
-          loc, val, ans->getValue().cast<ArrayAttr>(), success, failure);
+          loc, val, llvm::cast<ArrayAttr>(ans->getValue()), success, failure);
     else
       builder.create<pdl_interp::CheckTypeOp>(
-          loc, val, ans->getValue().cast<TypeAttr>(), success, failure);
+          loc, val, llvm::cast<TypeAttr>(ans->getValue()), success, failure);
     break;
   }
   case Predicates::AttributeQuestion: {
     auto *ans = cast<AttributeAnswer>(answer);
     builder.create<pdl_interp::CheckAttributeOp>(loc, val, ans->getValue(),
                                                  success, failure);
     break;
   }

mlir/lib/Conversion/TosaToLinalg/TosaToLinalg.cpp

createLinalgBodyCalculationForElementwiseOp(Operation *op, ValueRange args,
   if (isa<tosa::ExpOp>(op) && isa<FloatType>(elementTy))
     return rewriter.create<mlir::math::ExpOp>(loc, resultTypes, args);
 
   // tosa::TanhOp
   if (isa<tosa::TanhOp>(op) && isa<FloatType>(elementTy))
     return rewriter.create<mlir::math::TanhOp>(loc, resultTypes, args);
 
   // tosa::ErfOp
-  if (isa<tosa::ErfOp>(op) && elementTy.isa<FloatType>())
+  if (isa<tosa::ErfOp>(op) && llvm::isa<FloatType>(elementTy))
     return rewriter.create<mlir::math::ErfOp>(loc, resultTypes, args);
 
   // tosa::GreaterOp
   if (isa<tosa::GreaterOp>(op) && isa<FloatType>(elementTy))
     return rewriter.create<arith::CmpFOp>(loc, arith::CmpFPredicate::OGT,
                                           args[0], args[1]);
 
   if (isa<tosa::GreaterOp>(op) && elementTy.isSignlessInteger())
public:
   static llvm::SmallVector<Value> inferDynamicDimsForGather(OpBuilder &builder,
                                                             Location loc,
                                                             Value values,
                                                             Value indices) {
     llvm::SmallVector<Value> results;
 
     auto addDynamicDimension = [&](Value source, int64_t dim) {
       auto dynamicDim = tensor::createDimValue(builder, loc, source, dim);
-      if (auto dimValue = dynamicDim.value().dyn_cast<Value>())
+      if (auto dimValue = llvm::dyn_cast_if_present<Value>(dynamicDim.value()))
         results.push_back(dimValue);
     };
 
     addDynamicDimension(values, 0);
     addDynamicDimension(indices, 1);
     addDynamicDimension(values, 2);
     return results;
   }

mlir/lib/Debug/DebuggerExecutionContextHook.cpp

 
 void mlirDebuggerCursorSelectParentIRUnit() {
   auto &state = getGlobalDebuggerState();
   if (!state.cursor) {
     llvm::outs() << "No active MLIR cursor, select from the context first\n";
     return;
   }
   IRUnit *unit = &state.cursor;
-  if (auto *op = unit->dyn_cast<Operation *>()) {
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(*unit)) {
     state.cursor = op->getBlock();
-  } else if (auto *region = unit->dyn_cast<Region *>()) {
+  } else if (auto *region = llvm::dyn_cast_if_present<Region *>(*unit)) {
     state.cursor = region->getParentOp();
-  } else if (auto *block = unit->dyn_cast<Block *>()) {
+  } else if (auto *block = llvm::dyn_cast_if_present<Block *>(*unit)) {
     state.cursor = block->getParent();
   } else {
     llvm::outs() << "Current cursor is not a valid IRUnit";
     return;
   }
   state.cursor.print(llvm::outs());
   llvm::outs() << "\n";
 }
 
 void mlirDebuggerCursorSelectChildIRUnit(int index) {
   auto &state = getGlobalDebuggerState();
   if (!state.cursor) {
     llvm::outs() << "No active MLIR cursor, select from the context first\n";
     return;
   }
   IRUnit *unit = &state.cursor;
-  if (auto *op = unit->dyn_cast<Operation *>()) {
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(*unit)) {
     if (index < 0 || index >= static_cast<int>(op->getNumRegions())) {
       llvm::outs() << "Index invalid, op has " << op->getNumRegions()
                    << " but got " << index << "\n";
       return;
     }
     state.cursor = &op->getRegion(index);
-  } else if (auto *region = unit->dyn_cast<Region *>()) {
+  } else if (auto *region = llvm::dyn_cast_if_present<Region *>(*unit)) {
     auto block = region->begin();
     int count = 0;
     while (block != region->end() && count != index) {
       ++block;
       ++count;
     }
 
     if (block == region->end()) {
       llvm::outs() << "Index invalid, region has " << count << " block but got "
                    << index << "\n";
       return;
     }
     state.cursor = &*block;
-  } else if (auto *block = unit->dyn_cast<Block *>()) {
+  } else if (auto *block = llvm::dyn_cast_if_present<Block *>(*unit)) {
     auto op = block->begin();
     int count = 0;
     while (op != block->end() && count != index) {
       ++op;
       ++count;
     }
 
     if (op == block->end()) {

 
 void mlirDebuggerCursorSelectPreviousIRUnit() {
   auto &state = getGlobalDebuggerState();
   if (!state.cursor) {
     llvm::outs() << "No active MLIR cursor, select from the context first\n";
     return;
   }
   IRUnit *unit = &state.cursor;
-  if (auto *op = unit->dyn_cast<Operation *>()) {
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(*unit)) {
     Operation *previous = op->getPrevNode();
     if (!previous) {
       llvm::outs() << "No previous operation in the current block\n";
       return;
     }
     state.cursor = previous;
-  } else if (auto *region = unit->dyn_cast<Region *>()) {
+  } else if (auto *region = llvm::dyn_cast_if_present<Region *>(*unit)) {
     llvm::outs() << "Has region\n";
     Operation *parent = region->getParentOp();
     if (!parent) {
       llvm::outs() << "No parent operation for the current region\n";
       return;
     }
     if (region->getRegionNumber() == 0) {
       llvm::outs() << "No previous region in the current operation\n";
       return;
     }
     state.cursor =
         &region->getParentOp()->getRegion(region->getRegionNumber() - 1);
-  } else if (auto *block = unit->dyn_cast<Block *>()) {
+  } else if (auto *block = llvm::dyn_cast_if_present<Block *>(*unit)) {
     Block *previous = block->getPrevNode();
     if (!previous) {
       llvm::outs() << "No previous block in the current region\n";
       return;
     }
     state.cursor = previous;
   } else {
     llvm::outs() << "Current cursor is not a valid IRUnit";
     return;
   }
   state.cursor.print(llvm::outs());
   llvm::outs() << "\n";
 }
 
 void mlirDebuggerCursorSelectNextIRUnit() {
   auto &state = getGlobalDebuggerState();
   if (!state.cursor) {
     llvm::outs() << "No active MLIR cursor, select from the context first\n";
     return;
   }
   IRUnit *unit = &state.cursor;
-  if (auto *op = unit->dyn_cast<Operation *>()) {
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(*unit)) {
     Operation *next = op->getNextNode();
     if (!next) {
       llvm::outs() << "No next operation in the current block\n";
       return;
     }
     state.cursor = next;
-  } else if (auto *region = unit->dyn_cast<Region *>()) {
+  } else if (auto *region = llvm::dyn_cast_if_present<Region *>(*unit)) {
     Operation *parent = region->getParentOp();
     if (!parent) {
       llvm::outs() << "No parent operation for the current region\n";
       return;
     }
     if (region->getRegionNumber() == parent->getNumRegions() - 1) {
       llvm::outs() << "No next region in the current operation\n";
       return;
     }
     state.cursor =
         &region->getParentOp()->getRegion(region->getRegionNumber() + 1);
-  } else if (auto *block = unit->dyn_cast<Block *>()) {
+  } else if (auto *block = llvm::dyn_cast_if_present<Block *>(*unit)) {
     Block *next = block->getNextNode();
     if (!next) {
       llvm::outs() << "No next block in the current region\n";
       return;
     }
     state.cursor = next;
   } else {
     llvm::outs() << "Current cursor is not a valid IRUnit";

mlir/lib/Dialect/Affine/IR/AffineOps.cpp

static void materializeConstants(OpBuilder &b, Location loc,
   OpBuilder::Listener *listener = b.getListener();
   b.setListener(nullptr);
   auto listenerResetter =
       llvm::make_scope_exit([listener, &b] { b.setListener(listener); });
 
   actualValues.reserve(values.size());
   auto *dialect = b.getContext()->getLoadedDialect<AffineDialect>();
   for (OpFoldResult ofr : values) {
-    if (auto value = ofr.dyn_cast<Value>()) {
+    if (auto value = llvm::dyn_cast_if_present<Value>(ofr)) {
       actualValues.push_back(value);
       continue;
     }
     // Since we are directly specifying `index` as the result type, we need to
     // ensure the provided attribute is also an index type. Otherwise, the
     // AffineDialect materializer will create invalid `arith.constant`
     // operations if the provided Attribute is any other kind of integer.
     constants.push_back(dialect->materializeConstant(
void AffineDelinearizeIndexOp::build(OpBuilder &builder, OperationState &result,
   result.addTypes(SmallVector<Type>(basis.size(), builder.getIndexType()));
   result.addOperands(linearIndex);
   SmallVector<Value> basisValues =
       llvm::map_to_vector(basis, [&](OpFoldResult ofr) -> Value {
         std::optional<int64_t> staticDim = getConstantIntValue(ofr);
         if (staticDim.has_value())
           return builder.create<arith::ConstantIndexOp>(result.location,
                                                         *staticDim);
-        return ofr.dyn_cast<Value>();
+        return llvm::dyn_cast_if_present<Value>(ofr);
       });
   result.addOperands(basisValues);
 }
 
 LogicalResult AffineDelinearizeIndexOp::verify() {
   if (getBasis().empty())
     return emitOpError("basis should not be empty");
   if (getNumResults() != getBasis().size())

mlir/lib/Dialect/Arith/IR/ArithOps.cpp

 // OrIOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult arith::OrIOp::fold(FoldAdaptor adaptor) {
   /// or(x, 0) -> x
   if (matchPattern(getRhs(), m_Zero()))
     return getLhs();
   /// or(x, <all ones>) -> <all ones>
-  if (auto rhsAttr = adaptor.getRhs().dyn_cast_or_null<IntegerAttr>())
+  if (auto rhsAttr = llvm::dyn_cast_if_present<IntegerAttr>(adaptor.getRhs()))
     if (rhsAttr.getValue().isAllOnes())
       return rhsAttr;
 
   APInt intValue;
   /// or(x, xor(x, 1)) -> 1
   if (matchPattern(getRhs(), m_Op<XOrIOp>(matchers::m_Val(getLhs()),
                                           m_ConstantInt(&intValue))) &&
       intValue.isAllOnes())

 }
 
 //===----------------------------------------------------------------------===//
 // ExtFOp
 //===----------------------------------------------------------------------===//
 
 /// Always fold extension of FP constants.
 OpFoldResult arith::ExtFOp::fold(FoldAdaptor adaptor) {
-  auto constOperand = adaptor.getIn().dyn_cast_or_null<FloatAttr>();
+  auto constOperand = llvm::dyn_cast_if_present<FloatAttr>(adaptor.getIn());
   if (!constOperand)
     return {};
 
   // Convert to target type via 'double'.
   return FloatAttr::get(getType(), constOperand.getValue().convertToDouble());
 }
 
 bool arith::ExtFOp::areCastCompatible(TypeRange inputs, TypeRange outputs) {
for (auto pred : invPreds) {
         return getResult();
       }
     }
     llvm_unreachable("unknown cmpi predicate kind");
   }
 
   // We are moving constants to the right side; So if lhs is constant rhs is
   // guaranteed to be a constant.
-  if (auto lhs = adaptor.getLhs().dyn_cast_or_null<TypedAttr>()) {
+  if (auto lhs = llvm::dyn_cast_if_present<TypedAttr>(adaptor.getLhs())) {
     return constFoldBinaryOp<IntegerAttr>(
         adaptor.getOperands(), getI1SameShape(lhs.getType()),
         [pred = getPredicate()](const APInt &lhs, const APInt &rhs) {
           return APInt(1,
                        static_cast<int64_t>(applyCmpPredicate(pred, lhs, rhs)));
         });
   }
 
case arith::CmpFPredicate::UNO:
     return cmpResult == APFloat::cmpUnordered;
   case arith::CmpFPredicate::AlwaysTrue:
     return true;
   }
   llvm_unreachable("unknown cmpf predicate kind");
 }
 
 OpFoldResult arith::CmpFOp::fold(FoldAdaptor adaptor) {
-  auto lhs = adaptor.getLhs().dyn_cast_or_null<FloatAttr>();
-  auto rhs = adaptor.getRhs().dyn_cast_or_null<FloatAttr>();
+  auto lhs = llvm::dyn_cast_if_present<FloatAttr>(adaptor.getLhs());
+  auto rhs = llvm::dyn_cast_if_present<FloatAttr>(adaptor.getRhs());
 
   // If one operand is NaN, making them both NaN does not change the result.
   if (lhs && lhs.getValue().isNaN())
     rhs = lhs;
   if (rhs && rhs.getValue().isNaN())
     lhs = rhs;
 
   if (!lhs || !rhs)
if (pred == arith::CmpIPredicate::eq || pred == arith::CmpIPredicate::ne) {
           (cmpRhs == trueVal && cmpLhs == falseVal))
         return pred == arith::CmpIPredicate::ne ? trueVal : falseVal;
     }
   }
 
   // Constant-fold constant operands over non-splat constant condition.
   // select %cst_vec, %cst0, %cst1 => %cst2
   if (auto cond =
-          adaptor.getCondition().dyn_cast_or_null<DenseElementsAttr>()) {
+          llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getCondition())) {
     if (auto lhs =
-            adaptor.getTrueValue().dyn_cast_or_null<DenseElementsAttr>()) {
+            llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getTrueValue())) {
       if (auto rhs =
-              adaptor.getFalseValue().dyn_cast_or_null<DenseElementsAttr>()) {
+              llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getFalseValue())) {
         SmallVector<Attribute> results;
         results.reserve(static_cast<size_t>(cond.getNumElements()));
         auto condVals = llvm::make_range(cond.value_begin<BoolAttr>(),
                                          cond.value_end<BoolAttr>());
         auto lhsVals = llvm::make_range(lhs.value_begin<Attribute>(),
                                         lhs.value_end<Attribute>());
         auto rhsVals = llvm::make_range(rhs.value_begin<Attribute>(),
                                         rhs.value_end<Attribute>());

mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp

if (failed(trueType) || failed(falseType))
       return failure();
     if (*trueType == *falseType)
       return *trueType;
     if (trueType->getMemorySpace() != falseType->getMemorySpace())
       return op->emitError("inconsistent memory space on true/false operands");
 
     // If the buffers have different types, they differ only in their layout
     // map.
-    auto memrefType = trueType->cast<MemRefType>();
+    auto memrefType = llvm::cast<MemRefType>(*trueType);
     return getMemRefTypeWithFullyDynamicLayout(
         RankedTensorType::get(memrefType.getShape(),
                               memrefType.getElementType()),
         memrefType.getMemorySpace());
   }
 };
 
 } // namespace

mlir/lib/Dialect/Arith/Utils/Utils.cpp

 // Value. Returns `failure` when no folding happened.
 LogicalResult mlir::foldDynamicIndexList(Builder &b,
                                          SmallVectorImpl<OpFoldResult> &ofrs) {
   bool valuesChanged = false;
   for (OpFoldResult &ofr : ofrs) {
     if (ofr.is<Attribute>())
       continue;
     // Newly static, move from Value to constant.
-    if (auto cstOp =
-            ofr.dyn_cast<Value>().getDefiningOp<arith::ConstantIndexOp>()) {
+    if (auto cstOp = llvm::dyn_cast_if_present<Value>(ofr)
+                         .getDefiningOp<arith::ConstantIndexOp>()) {
       ofr = b.getIndexAttr(cstOp.value());
       valuesChanged = true;
     }
   }
   return success(valuesChanged);
 }
 
 llvm::SmallBitVector mlir::getPositionsOfShapeOne(unsigned rank,
                                                   ArrayRef<int64_t> shape) {
   llvm::SmallBitVector dimsToProject(shape.size());
   for (unsigned pos = 0, e = shape.size(); pos < e && rank > 0; ++pos) {
     if (shape[pos] == 1) {
       dimsToProject.set(pos);
       --rank;
     }
   }
   return dimsToProject;
 }
 
 Value mlir::getValueOrCreateConstantIndexOp(OpBuilder &b, Location loc,
                                             OpFoldResult ofr) {
-  if (auto value = ofr.dyn_cast<Value>())
+  if (auto value = llvm::dyn_cast_if_present<Value>(ofr))
     return value;
-  auto attr = dyn_cast<IntegerAttr>(ofr.dyn_cast<Attribute>());
+  auto attr = dyn_cast<IntegerAttr>(llvm::dyn_cast_if_present<Attribute>(ofr));
   assert(attr && "expect the op fold result casts to an integer attribute");
   return b.create<arith::ConstantIndexOp>(loc, attr.getValue().getSExtValue());
 }
 
 Value mlir::getValueOrCreateCastToIndexLike(OpBuilder &b, Location loc,
                                             Type targetType, Value value) {
   if (targetType == value.getType())
     return value;

mlir/lib/Dialect/Bufferization/IR/BufferizationOps.cpp

LogicalResult AllocTensorOp::bufferize(RewriterBase &rewriter,
   if (failed(allocType))
     return failure();
   SmallVector<Value> dynamicDims = getDynamicSizes();
   if (getCopy()) {
     assert(dynamicDims.empty() && "expected either `copy` or `dynamicDims`");
     populateDynamicDimSizes(rewriter, loc, copyBuffer, dynamicDims);
   }
   FailureOr<Value> alloc = options.createAlloc(
-      rewriter, loc, allocType->cast<MemRefType>(), dynamicDims);
+      rewriter, loc, llvm::cast<MemRefType>(*allocType), dynamicDims);
   if (failed(alloc))
     return failure();
 
   // Create memory copy (if any).
   if (getCopy()) {
     if (failed(options.createMemCpy(rewriter, loc, copyBuffer, *alloc)))
       return failure();
   }

mlir/lib/Dialect/Bufferization/Transforms/DropEquivalentBufferResults.cpp

if (auto candidateOp = dyn_cast<func::ReturnOp>(b.getTerminator())) {
       returnOp = candidateOp;
     }
   }
   return returnOp;
 }
 
 /// Return the func::FuncOp called by `callOp`.
 static func::FuncOp getCalledFunction(CallOpInterface callOp) {
-  SymbolRefAttr sym = callOp.getCallableForCallee().dyn_cast<SymbolRefAttr>();
+  SymbolRefAttr sym =
+      llvm::dyn_cast_if_present<SymbolRefAttr>(callOp.getCallableForCallee());
   if (!sym)
     return nullptr;
   return dyn_cast_or_null<func::FuncOp>(
       SymbolTable::lookupNearestSymbolFrom(callOp, sym));
 }
 
 LogicalResult
 mlir::bufferization::dropEquivalentBufferResults(ModuleOp module) {

mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp

getBufferizedFunctionArgType(FuncOp funcOp, int64_t index,
   assert(rankedMemrefType && "buffer layout not supported on unranked tensors");
   return MemRefType::get(
       rankedMemrefType.getShape(), rankedMemrefType.getElementType(),
       layoutAttr.getValue(), rankedMemrefType.getMemorySpace());
 }
 
 /// Return the FuncOp called by `callOp`.
 static FuncOp getCalledFunction(CallOpInterface callOp) {
-  SymbolRefAttr sym = callOp.getCallableForCallee().dyn_cast<SymbolRefAttr>();
+  SymbolRefAttr sym = llvm::dyn_cast_if_present<SymbolRefAttr>(callOp.getCallableForCallee());
   if (!sym)
     return nullptr;
   return dyn_cast_or_null<FuncOp>(
       SymbolTable::lookupNearestSymbolFrom(callOp, sym));
 }
 
 /// Get FuncAnalysisState.
 static const FuncAnalysisState &

mlir/lib/Dialect/Bufferization/Transforms/OneShotAnalysis.cpp

 
 static void annotateOpsWithAliasSets(Operation *op,
                                      const OneShotAnalysisState &state) {
   AsmState asmState(op);
   Builder b(op->getContext());
   op->walk([&](Operation *op) {
     SmallVector<Attribute> aliasSets;
     for (OpResult opResult : op->getOpResults()) {
-      if (opResult.getType().isa<TensorType>()) {
+      if (llvm::isa<TensorType>(opResult.getType())) {
         SmallVector<Attribute> aliases;
         state.applyOnAliases(opResult, [&](Value alias) {
           std::string buffer;
           llvm::raw_string_ostream stream(buffer);
           alias.printAsOperand(stream, asmState);
           aliases.push_back(b.getStringAttr(stream.str()));
         });
         aliasSets.push_back(b.getArrayAttr(aliases));

mlir/lib/Dialect/Bufferization/Transforms/OneShotModuleBufferize.cpp

static void removeBufferizationAttributes(BlockArgument bbArg) {
   funcOp.removeArgAttr(bbArg.getArgNumber(),
                        BufferizationDialect::kBufferLayoutAttrName);
   funcOp.removeArgAttr(bbArg.getArgNumber(),
                        BufferizationDialect::kWritableAttrName);
 }
 
 /// Return the func::FuncOp called by `callOp`.
 static func::FuncOp getCalledFunction(func::CallOp callOp) {
-  SymbolRefAttr sym = callOp.getCallableForCallee().dyn_cast<SymbolRefAttr>();
+  SymbolRefAttr sym = llvm::dyn_cast_if_present<SymbolRefAttr>(callOp.getCallableForCallee());
   if (!sym)
     return nullptr;
   return dyn_cast_or_null<func::FuncOp>(
       SymbolTable::lookupNearestSymbolFrom(callOp, sym));
 }
 
 /// Gather equivalence info of CallOps.
 /// Note: This only adds new equivalence info if the called function was already

mlir/lib/Dialect/Complex/IR/ComplexOps.cpp

OpFoldResult CreateOp::fold(FoldAdaptor adaptor) {
   return {};
 }
 
 //===----------------------------------------------------------------------===//
 // ImOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult ImOp::fold(FoldAdaptor adaptor) {
-  ArrayAttr arrayAttr = adaptor.getComplex().dyn_cast_or_null<ArrayAttr>();
+  ArrayAttr arrayAttr =
+      llvm::dyn_cast_if_present<ArrayAttr>(adaptor.getComplex());
   if (arrayAttr && arrayAttr.size() == 2)
     return arrayAttr[1];
   if (auto createOp = getOperand().getDefiningOp<CreateOp>())
     return createOp.getOperand(1);
   return {};
 }
 
 //===----------------------------------------------------------------------===//
 // ReOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult ReOp::fold(FoldAdaptor adaptor) {
-  ArrayAttr arrayAttr = adaptor.getComplex().dyn_cast_or_null<ArrayAttr>();
+  ArrayAttr arrayAttr =
+      llvm::dyn_cast_if_present<ArrayAttr>(adaptor.getComplex());
   if (arrayAttr && arrayAttr.size() == 2)
     return arrayAttr[0];
   if (auto createOp = getOperand().getDefiningOp<CreateOp>())
     return createOp.getOperand(0);
   return {};
 }
 
 //===----------------------------------------------------------------------===//

mlir/lib/Dialect/DLTI/DLTI.cpp

if (failed(parser.parseComma()) || failed(parser.parseAttribute(value)) ||
     return {};
 
   return type ? get(type, value)
               : get(parser.getBuilder().getStringAttr(identifier), value);
 }
 
 void DataLayoutEntryAttr::print(AsmPrinter &os) const {
   os << DataLayoutEntryAttr::kAttrKeyword << "<";
-  if (auto type = getKey().dyn_cast<Type>())
+  if (auto type = llvm::dyn_cast_if_present<Type>(getKey()))
     os << type;
   else
     os << "\"" << getKey().get<StringAttr>().strref() << "\"";
   os << ", " << getValue() << ">";
 }
 
 //===----------------------------------------------------------------------===//
 // DataLayoutSpecAttr

 }
 
 LogicalResult
 DataLayoutSpecAttr::verify(function_ref<InFlightDiagnostic()> emitError,
                            ArrayRef<DataLayoutEntryInterface> entries) {
   DenseSet<Type> types;
   DenseSet<StringAttr> ids;
   for (DataLayoutEntryInterface entry : entries) {
-    if (auto type = entry.getKey().dyn_cast<Type>()) {
+    if (auto type = llvm::dyn_cast_if_present<Type>(entry.getKey())) {
       if (!types.insert(type).second)
         return emitError() << "repeated layout entry key: " << type;
     } else {
       auto id = entry.getKey().get<StringAttr>();
       if (!ids.insert(id).second)
         return emitError() << "repeated layout entry key: " << id.getValue();
     }
   }

mlir/lib/Dialect/LLVMIR/IR/LLVMDialect.cpp

static void destructureIndices(Type currType, ArrayRef<GEPArg> indices,
                                SmallVectorImpl<Value> &dynamicIndices) {
   for (const GEPArg &iter : indices) {
     // If the thing we are currently indexing into is a struct we must turn
     // any integer constants into constant indices. If this is not possible
     // we don't do anything here. The verifier will catch it and emit a proper
     // error. All other canonicalization is done in the fold method.
     bool requiresConst = !rawConstantIndices.empty() &&
                          currType.isa_and_nonnull<LLVMStructType>();
-    if (Value val = iter.dyn_cast<Value>()) {
+    if (Value val = llvm::dyn_cast_if_present<Value>(iter)) {
       APInt intC;
       if (requiresConst && matchPattern(val, m_ConstantInt(&intC)) &&
           intC.isSignedIntN(kGEPConstantBitWidth)) {
         rawConstantIndices.push_back(intC.getSExtValue());
       } else {
         rawConstantIndices.push_back(GEPOp::kDynamicIndex);
         dynamicIndices.push_back(val);
       }

 }
 
 static void printGEPIndices(OpAsmPrinter &printer, LLVM::GEPOp gepOp,
                             OperandRange indices,
                             DenseI32ArrayAttr rawConstantIndices) {
   llvm::interleaveComma(
       GEPIndicesAdaptor<OperandRange>(rawConstantIndices, indices), printer,
       [&](PointerUnion<IntegerAttr, Value> cst) {
-        if (Value val = cst.dyn_cast<Value>())
+        if (Value val = llvm::dyn_cast_if_present<Value>(cst))
           printer.printOperand(val);
         else
           printer << cst.get<IntegerAttr>().getInt();
       });
 }
 
 namespace {
 /// Base class for llvm::Error related to GEP index.
OpFoldResult LLVM::GEPOp::fold(FoldAdaptor adaptor) {
   for (auto iter : llvm::enumerate(indices)) {
     auto integer = llvm::dyn_cast_or_null<IntegerAttr>(iter.value());
     // Constant indices can only be int32_t, so if integer does not fit we
     // are forced to keep it dynamic, despite being a constant.
     if (!indices.isDynamicIndex(iter.index()) || !integer ||
         !integer.getValue().isSignedIntN(kGEPConstantBitWidth)) {
 
       PointerUnion<IntegerAttr, Value> existing = getIndices()[iter.index()];
-      if (Value val = existing.dyn_cast<Value>())
+      if (Value val = llvm::dyn_cast_if_present<Value>(existing))
         gepArgs.emplace_back(val);
       else
         gepArgs.emplace_back(existing.get<IntegerAttr>().getInt());
 
       continue;
     }
 
     changed = true;

mlir/lib/Dialect/LLVMIR/IR/LLVMMemorySlot.cpp

 }
 
 //===----------------------------------------------------------------------===//
 // Interfaces for GEPOp
 //===----------------------------------------------------------------------===//
 
 static bool hasAllZeroIndices(LLVM::GEPOp gepOp) {
   return llvm::all_of(gepOp.getIndices(), [](auto index) {
-    auto indexAttr = index.template dyn_cast<IntegerAttr>();
+    auto indexAttr = llvm::dyn_cast_if_present<IntegerAttr>(index);
     return indexAttr && indexAttr.getValue() == 0;
   });
 }
 
 bool LLVM::GEPOp::canUsesBeRemoved(
     const SmallPtrSetImpl<OpOperand *> &blockingUses,
     SmallVectorImpl<OpOperand *> &newBlockingUses) {
   // GEP can be removed as long as it is a no-op and its users can be removed.

 /// is not clear, returns null type.
 static Type computeReachedGEPType(LLVM::GEPOp gep) {
   if (gep.getIndices().empty())
     return {};
 
   // Ensures all indices are static and fetches them.
   SmallVector<IntegerAttr> indices;
   for (auto index : gep.getIndices()) {
-    IntegerAttr indexInt = index.dyn_cast<IntegerAttr>();
+    IntegerAttr indexInt = llvm::dyn_cast_if_present<IntegerAttr>(index);
     if (!indexInt)
       return {};
     indices.push_back(indexInt);
   }
 
   // Check the pointer indexing only targets the first element.
   if (indices[0].getInt() != 0)
     return {};
 
   // Set the initial type currently being used for indexing. This will be
   // updated as the indices get walked over.
   std::optional<Type> maybeSelectedType = gep.getElemType();
   if (!maybeSelectedType)
     return {};
   Type selectedType = *maybeSelectedType;
 
   // Follow the indexed elements in the gep.
   for (IntegerAttr index : llvm::drop_begin(indices)) {
     // Ensure the structure of the type being indexed can be reasoned about.
     // This includes rejecting any potential typed pointer.
-    auto destructurable = selectedType.dyn_cast<DestructurableTypeInterface>();
+    auto destructurable = llvm::dyn_cast<DestructurableTypeInterface>(selectedType);
     if (!destructurable)
       return {};
 
     // Follow the type at the index the gep is accessing, making it the new type
     // used for indexing.
     Type field = destructurable.getTypeAtIndex(index);
     if (!field)
       return {};
if (!reachedType)
     return failure();
   mustBeSafelyUsed.emplace_back<MemorySlot>({getResult(), reachedType});
   return success();
 }
 
 bool LLVM::GEPOp::canRewire(const DestructurableMemorySlot &slot,
                             SmallPtrSetImpl<Attribute> &usedIndices,
                             SmallVectorImpl<MemorySlot> &mustBeSafelyUsed) {
-  auto basePtrType = getBase().getType().dyn_cast<LLVM::LLVMPointerType>();
+  auto basePtrType = llvm::dyn_cast<LLVM::LLVMPointerType>(getBase().getType());
   if (!basePtrType)
     return false;
 
   // Typed pointers are not supported. This should be removed once typed
   // pointers are removed from the LLVM dialect.
   if (!basePtrType.isOpaque())
     return false;
 
   if (getBase() != slot.ptr || slot.elemType != getElemType())
     return false;
   Type reachedType = computeReachedGEPType(*this);
   if (!reachedType || getIndices().size() < 2)
     return false;
   auto firstLevelIndex = cast<IntegerAttr>(getIndices()[1]);
   assert(slot.elementPtrs.contains(firstLevelIndex));
-  if (!slot.elementPtrs.at(firstLevelIndex).isa<LLVM::LLVMPointerType>())
+  if (!llvm::isa<LLVM::LLVMPointerType>(slot.elementPtrs.at(firstLevelIndex)))
     return false;
   mustBeSafelyUsed.emplace_back<MemorySlot>({getResult(), reachedType});
   usedIndices.insert(firstLevelIndex);
   return true;
 }
 
 DeletionKind LLVM::GEPOp::rewire(const DestructurableMemorySlot &slot,
                                  DenseMap<Attribute, MemorySlot> &subslots,
                                  RewriterBase &rewriter) {
-  IntegerAttr firstLevelIndex = getIndices()[1].dyn_cast<IntegerAttr>();
+  IntegerAttr firstLevelIndex = llvm::dyn_cast_if_present<IntegerAttr>(getIndices()[1]);
   const MemorySlot &newSlot = subslots.at(firstLevelIndex);
 
   ArrayRef<int32_t> remainingIndices = getRawConstantIndices().slice(2);
 
   // If the GEP would become trivial after this transformation, eliminate it.
   // A GEP should only be eliminated if it has no indices (except the first
   // pointer index), as simplifying GEPs with all-zero indices would eliminate
   // structure information useful for further destruction.
LLVM::LLVMStructType::getSubelementIndexMap() {
   Type i32 = IntegerType::get(getContext(), 32);
   DenseMap<Attribute, Type> destructured;
   for (const auto &[index, elemType] : llvm::enumerate(getBody()))
     destructured.insert({IntegerAttr::get(i32, index), elemType});
   return destructured;
 }
 
 Type LLVM::LLVMStructType::getTypeAtIndex(Attribute index) {
-  auto indexAttr = index.dyn_cast<IntegerAttr>();
+  auto indexAttr = llvm::dyn_cast<IntegerAttr>(index);
   if (!indexAttr || !indexAttr.getType().isInteger(32))
     return {};
   int32_t indexInt = indexAttr.getInt();
   ArrayRef<Type> body = getBody();
   if (indexInt < 0 || body.size() <= static_cast<uint32_t>(indexInt))
     return {};
   return body[indexInt];
 }
 
 std::optional<DenseMap<Attribute, Type>>
 LLVM::LLVMArrayType::getSubelementIndexMap() const {
   constexpr size_t maxArraySizeForDestructuring = 16;
   if (getNumElements() > maxArraySizeForDestructuring)
     return {};
   int32_t numElements = getNumElements();
 
   Type i32 = IntegerType::get(getContext(), 32);
   DenseMap<Attribute, Type> destructured;
   for (int32_t index = 0; index < numElements; ++index)
     destructured.insert({IntegerAttr::get(i32, index), getElementType()});
   return destructured;
 }
 
 Type LLVM::LLVMArrayType::getTypeAtIndex(Attribute index) const {
-  auto indexAttr = index.dyn_cast<IntegerAttr>();
+  auto indexAttr = llvm::dyn_cast<IntegerAttr>(index);
   if (!indexAttr || !indexAttr.getType().isInteger(32))
     return {};
   int32_t indexInt = indexAttr.getInt();
   if (indexInt < 0 || getNumElements() <= static_cast<uint32_t>(indexInt))
     return {};
   return getElementType();
 }

mlir/lib/Dialect/LLVMIR/IR/LLVMTypes.cpp

bool LLVMPointerType::areCompatible(DataLayoutEntryListRef oldLayout,
   for (DataLayoutEntryInterface newEntry : newLayout) {
     if (!newEntry.isTypeEntry())
       continue;
     unsigned size = kDefaultPointerSizeBits;
     unsigned abi = kDefaultPointerAlignment;
     auto newType = llvm::cast<LLVMPointerType>(newEntry.getKey().get<Type>());
     const auto *it =
         llvm::find_if(oldLayout, [&](DataLayoutEntryInterface entry) {
-          if (auto type = entry.getKey().dyn_cast<Type>()) {
+          if (auto type = llvm::dyn_cast_if_present<Type>(entry.getKey())) {
             return llvm::cast<LLVMPointerType>(type).getAddressSpace() ==
                    newType.getAddressSpace();
           }
           return false;
         });
     if (it == oldLayout.end()) {
       llvm::find_if(oldLayout, [&](DataLayoutEntryInterface entry) {
-        if (auto type = entry.getKey().dyn_cast<Type>()) {
+        if (auto type = llvm::dyn_cast_if_present<Type>(entry.getKey())) {
           return llvm::cast<LLVMPointerType>(type).getAddressSpace() == 0;
         }
         return false;
       });
     }
     if (it != oldLayout.end()) {
       size = *extractPointerSpecValue(*it, PtrDLEntryPos::Size);
       abi = *extractPointerSpecValue(*it, PtrDLEntryPos::Abi);

mlir/lib/Dialect/Linalg/TransformOps/LinalgTransformOps.cpp

for (auto [i, op] : llvm::enumerate(targets)) {
     scf::SCFTilingOptions tilingOptions;
     if (!tileSizes.empty()) {
       tilingOptions.setTileSizeComputationFunction([&, index = i](OpBuilder &b,
                                                                   Operation *) {
         SmallVector<Value, 4> sizes;
         sizes.reserve(tileSizes.size());
         unsigned dynamicIdx = 0;
         for (OpFoldResult ofr : getMixedSizes()) {
-          if (auto attr = ofr.dyn_cast<Attribute>()) {
+          if (auto attr = llvm::dyn_cast_if_present<Attribute>(ofr)) {
             sizes.push_back(b.create<arith::ConstantIndexOp>(
                 getLoc(), cast<IntegerAttr>(attr).getInt()));
             continue;
           }
           ArrayRef<Operation *> dynamicSizes = dynamicSizeProducers[dynamicIdx];
           ArrayRef<int64_t> params = paramSizes[dynamicIdx];
           ++dynamicIdx;
           assert((dynamicSizes.empty() ^ params.empty()) &&
for (auto en : llvm::enumerate(targets)) {
     unsigned index = en.index();
     if (!tileSizes.empty()) {
       tilingOptions.setTileSizeComputationFunction(
           [&, index](OpBuilder &b, Operation *) {
             SmallVector<Value, 4> sizes;
             sizes.reserve(tileSizes.size());
             unsigned dynamicIdx = 0;
             for (OpFoldResult ofr : getMixedSizes()) {
-              if (auto attr = ofr.dyn_cast<Attribute>()) {
+              if (auto attr = llvm::dyn_cast_if_present<Attribute>(ofr)) {
                 sizes.push_back(b.create<arith::ConstantIndexOp>(
                     getLoc(), cast<IntegerAttr>(attr).getInt()));
               } else {
                 sizes.push_back(
                     dynamicSizeProducers[dynamicIdx++][index]->getResult(0));
               }
             }
             return sizes;

mlir/lib/Dialect/Linalg/Transforms/ElementwiseOpFusion.cpp

return rewriter.notifyMatchFailure(
         genericOp, "illegal to collapse specified dimensions");
   }
 
   // Bail on non-canonical ranges.
   SmallVector<Range> loopRanges =
       cast<LinalgOp>(genericOp.getOperation())
           .createLoopRanges(rewriter, genericOp.getLoc());
   auto opFoldIsConstantValue = [](OpFoldResult ofr, int64_t value) {
-    if (auto attr = ofr.dyn_cast<Attribute>())
+    if (auto attr = llvm::dyn_cast_if_present<Attribute>(ofr))
       return cast<IntegerAttr>(attr).getInt() == value;
     llvm::APInt actual;
     return matchPattern(ofr.get<Value>(), m_ConstantInt(&actual)) &&
            actual.getSExtValue() == value;
   };
   if (!llvm::all_of(loopRanges, [&](Range range) {
         return opFoldIsConstantValue(range.offset, 0) &&
                opFoldIsConstantValue(range.stride, 1);

mlir/lib/Dialect/Linalg/Transforms/Promotion.cpp

FailureOr<PromotionInfo> mlir::linalg::promoteSubviewAsNewBuffer(
   for (const auto &en : llvm::enumerate(subView.getOrCreateRanges(b, loc))) {
     if (droppedDims[en.index()])
       continue;
     auto rangeValue = en.value();
     // Try to extract a tight constant. If the size is known statically, no need
     // to look for the bound.
     LLVM_DEBUG(llvm::dbgs() << "Extract tightest: " << rangeValue.size << "\n");
     Value size;
-    if (auto attr = rangeValue.size.dyn_cast<Attribute>()) {
+    if (auto attr = llvm::dyn_cast_if_present<Attribute>(rangeValue.size)) {
       size = getValueOrCreateConstantIndexOp(b, loc, rangeValue.size);
     } else {
       Value materializedSize =
           getValueOrCreateConstantIndexOp(b, loc, rangeValue.size);
       FailureOr<int64_t> upperBound =
           ValueBoundsConstraintSet::computeConstantBound(
               presburger::BoundType::UB, materializedSize, /*dim=*/std::nullopt,
               /*stopCondition=*/nullptr, /*closedUB=*/true);

mlir/lib/Dialect/Linalg/Transforms/Split.cpp

OpFoldResult minSplitPoint = affine::makeComposedFoldedAffineMin(
       {splitPoint, offsets[dimension], sizes[dimension]});
 
   // Compute the size of the second part. Return early if the second part would
   // have an empty iteration space.
   OpFoldResult remainingSize = affine::makeComposedFoldedAffineApply(
       rewriter, op.getLoc(), d0 + d1 - d2,
       {iterationSpace[dimension].offset, iterationSpace[dimension].size,
        minSplitPoint});
-  if (auto attr = remainingSize.dyn_cast<Attribute>()) {
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(remainingSize)) {
     if (cast<IntegerAttr>(attr).getValue().isZero())
       return {op, TilingInterface()};
   }
 
   // Compute destination tensors.
   SmallVector<Value> destinationTensors;
   LogicalResult destStatus = tensor::getOrCreateDestinations(
       rewriter, op.getLoc(), op, destinationTensors);

mlir/lib/Dialect/Linalg/Transforms/Tiling.cpp

 using namespace mlir::linalg;
 using namespace mlir::scf;
 
 #define DEBUG_TYPE "linalg-tiling"
 
 static bool isZero(OpFoldResult v) {
   if (!v)
     return false;
-  if (auto attr = v.dyn_cast<Attribute>()) {
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(v)) {
     IntegerAttr intAttr = dyn_cast<IntegerAttr>(attr);
     return intAttr && intAttr.getValue().isZero();
   }
   if (auto cst = v.get<Value>().getDefiningOp<arith::ConstantIndexOp>())
     return cst.value() == 0;
   return false;
 }
 
void mlir::linalg::transformIndexOps(
   offsetIndices(b, op, getAsOpFoldResult(allIvs));
 }
 
 /// Asserts that the given index-typed value is strictly positive. If the value
 /// is an attribute, asserts at compile time, otherwise emits an assertion
 /// checked at runtime.
 static void emitIsPositiveIndexAssertion(ImplicitLocOpBuilder &b,
                                          OpFoldResult value) {
-  if (auto attr = value.dyn_cast<Attribute>()) {
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(value)) {
     assert(cast<IntegerAttr>(attr).getValue().isStrictlyPositive() &&
            "expected strictly positive tile size and divisor");
     return;
   }
 
   Value zero = b.create<arith::ConstantIndexOp>(0);
   Value condition = b.create<arith::CmpIOp>(arith::CmpIPredicate::sgt,
                                             value.get<Value>(), zero);

mlir/lib/Dialect/Linalg/Transforms/Transforms.cpp

Value GeneralizePadOpPattern::createFillOrGenerateOp(
   return generateOp;
 }
 
 LogicalResult
 GeneralizePadOpPattern::matchAndRewrite(tensor::PadOp padOp,
                                         PatternRewriter &rewriter) const {
   // Given an OpFoldResult, return an index-typed value.
   auto getIdxValue = [&](OpFoldResult ofr) {
-    if (auto val = ofr.dyn_cast<Value>())
+    if (auto val = llvm::dyn_cast_if_present<Value>(ofr))
       return val;
     return rewriter
         .create<arith::ConstantIndexOp>(
             padOp.getLoc(), cast<IntegerAttr>(ofr.get<Attribute>()).getInt())
         .getResult();
   };
 
   auto resultType = padOp.getResultType();

mlir/lib/Dialect/Linalg/Transforms/Vectorization.cpp

 
 /// Given an ArrayRef of OpFoldResults, return a vector of Values.
 /// IntegerAttrs are converted to ConstantIndexOps. Other attribute types are
 /// not supported.
 static SmallVector<Value> ofrToIndexValues(RewriterBase &rewriter, Location loc,
                                            ArrayRef<OpFoldResult> ofrs) {
   SmallVector<Value> result;
   for (auto o : ofrs) {
-    if (auto val = o.template dyn_cast<Value>()) {
+    if (auto val = llvm::dyn_cast_if_present<Value>(o)) {
       result.push_back(val);
     } else {
       result.push_back(rewriter.create<arith::ConstantIndexOp>(
           loc, getIntFromAttr(o.template get<Attribute>())));
     }
   }
   return result;
 }
for (unsigned i = 0; i < t1.getRank(); ++i) {
       auto size1 = beforeSlice.getMixedSizes()[i];
       auto size2 = afterTrimming.getMixedSizes()[i];
 
       // Case 1: Same value or same constant int.
       if (isEqualConstantIntOrValue(size1, size2))
         continue;
 
       // Other cases: Take a deeper look at defining ops of values.
-      auto v1 = size1.dyn_cast<Value>();
-      auto v2 = size2.dyn_cast<Value>();
+      auto v1 = llvm::dyn_cast_if_present<Value>(size1);
+      auto v2 = llvm::dyn_cast_if_present<Value>(size2);
       if (!v1 || !v2)
         return false;
 
       // Case 2: Both values are identical AffineMinOps. (Should not happen if
       // CSE is run.)
       auto minOp1 = v1.getDefiningOp<affine::AffineMinOp>();
       auto minOp2 = v2.getDefiningOp<affine::AffineMinOp>();
       if (minOp1 && minOp2 && minOp1.getAffineMap() == minOp2.getAffineMap() &&

mlir/lib/Dialect/Linalg/Utils/Utils.cpp

 std::optional<SmallVector<ReassociationIndices>>
 getReassociationMapForFoldingUnitDims(ArrayRef<OpFoldResult> mixedSizes) {
   SmallVector<ReassociationIndices> reassociation;
   ReassociationIndices curr;
   for (const auto &it : llvm::enumerate(mixedSizes)) {
     auto dim = it.index();
     auto size = it.value();
     curr.push_back(dim);
-    auto attr = size.dyn_cast<Attribute>();
+    auto attr = llvm::dyn_cast_if_present<Attribute>(size);
     if (attr && cast<IntegerAttr>(attr).getInt() == 1)
       continue;
     reassociation.emplace_back(ReassociationIndices{});
     std::swap(reassociation.back(), curr);
   }
   // When the reassociations are not empty, then fold the remaining
   // unit-dimensions into the last dimension.  If the reassociations so far is
   // empty, then leave it emtpy. This will fold everything to a rank-0 tensor.

mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp

static void walkIndicesAsAttr(MLIRContext *ctx, ArrayRef<int64_t> shape,
   } while (succeeded(nextIndex(shape, shapeIter)));
 }
 
 //===----------------------------------------------------------------------===//
 //  Interfaces for AllocaOp
 //===----------------------------------------------------------------------===//
 
 static bool isSupportedElementType(Type type) {
-  return type.isa<MemRefType>() ||
+  return llvm::isa<MemRefType>(type) ||
          OpBuilder(type.getContext()).getZeroAttr(type);
 }
 
 SmallVector<MemorySlot> memref::AllocaOp::getPromotableSlots() {
   MemRefType type = getType();
   if (!isSupportedElementType(type.getElementType()))
     return {};
   if (!type.hasStaticShape())

 
 void memref::AllocaOp::handleBlockArgument(const MemorySlot &slot,
                                            BlockArgument argument,
                                            RewriterBase &rewriter) {}
 
 SmallVector<DestructurableMemorySlot>
 memref::AllocaOp::getDestructurableSlots() {
   MemRefType memrefType = getType();
-  auto destructurable = memrefType.dyn_cast<DestructurableTypeInterface>();
+  auto destructurable = llvm::dyn_cast<DestructurableTypeInterface>(memrefType);
   if (!destructurable)
     return {};
 
   Optional<DenseMap<Attribute, Type>> destructuredType =
       destructurable.getSubelementIndexMap();
   if (!destructuredType)
     return {};
 
   DenseMap<Attribute, Type> indexMap;
   for (auto const &[index, type] : *destructuredType)
     indexMap.insert({index, MemRefType::get({}, type)});
 
   return {DestructurableMemorySlot{{getMemref(), memrefType}, indexMap}};
 }
 
 DenseMap<Attribute, MemorySlot>
 memref::AllocaOp::destructure(const DestructurableMemorySlot &slot,
                               const SmallPtrSetImpl<Attribute> &usedIndices,
                               RewriterBase &rewriter) {
   rewriter.setInsertionPointAfter(*this);
 
   DenseMap<Attribute, MemorySlot> slotMap;
 
-  auto memrefType = getType().cast<DestructurableTypeInterface>();
+  auto memrefType = llvm::cast<DestructurableTypeInterface>(getType());
   for (Attribute usedIndex : usedIndices) {
     Type elemType = memrefType.getTypeAtIndex(usedIndex);
     MemRefType elemPtr = MemRefType::get({}, elemType);
     auto subAlloca = rewriter.create<memref::AllocaOp>(getLoc(), elemPtr);
     slotMap.try_emplace<MemorySlot>(usedIndex,
                                     {subAlloca.getResult(), elemType});
   }
 

 
 namespace {
 
 struct MemRefDestructurableTypeExternalModel
     : public DestructurableTypeInterface::ExternalModel<
           MemRefDestructurableTypeExternalModel, MemRefType> {
   std::optional<DenseMap<Attribute, Type>>
   getSubelementIndexMap(Type type) const {
-    auto memrefType = type.cast<MemRefType>();
+    auto memrefType = llvm::cast<MemRefType>(type);
     constexpr int64_t maxMemrefSizeForDestructuring = 16;
     if (!memrefType.hasStaticShape() ||
         memrefType.getNumElements() > maxMemrefSizeForDestructuring ||
         memrefType.getNumElements() == 1)
       return {};
 
     DenseMap<Attribute, Type> destructured;
     walkIndicesAsAttr(
         memrefType.getContext(), memrefType.getShape(), [&](Attribute index) {
           destructured.insert({index, memrefType.getElementType()});
         });
 
     return destructured;
   }
 
   Type getTypeAtIndex(Type type, Attribute index) const {
-    auto memrefType = type.cast<MemRefType>();
-    auto coordArrAttr = index.dyn_cast<ArrayAttr>();
+    auto memrefType = llvm::cast<MemRefType>(type);
+    auto coordArrAttr = llvm::dyn_cast<ArrayAttr>(index);
     if (!coordArrAttr || coordArrAttr.size() != memrefType.getShape().size())
       return {};
 
     Type indexType = IndexType::get(memrefType.getContext());
     for (const auto &[coordAttr, dimSize] :
          llvm::zip(coordArrAttr, memrefType.getShape())) {
-      auto coord = coordAttr.dyn_cast<IntegerAttr>();
+      auto coord = llvm::dyn_cast<IntegerAttr>(coordAttr);
       if (!coord || coord.getType() != indexType || coord.getInt() < 0 ||
           coord.getInt() >= dimSize)
         return {};
     }
 
     return memrefType.getElementType();
   }
 };

mlir/lib/Dialect/MemRef/IR/MemRefOps.cpp

 static std::optional<llvm::SmallBitVector>
 computeMemRefRankReductionMask(MemRefType originalType, MemRefType reducedType,
                                ArrayRef<OpFoldResult> sizes) {
   llvm::SmallBitVector unusedDims(originalType.getRank());
   if (originalType.getRank() == reducedType.getRank())
     return unusedDims;
 
   for (const auto &dim : llvm::enumerate(sizes))
-    if (auto attr = dim.value().dyn_cast<Attribute>())
+    if (auto attr = llvm::dyn_cast_if_present<Attribute>(dim.value()))
       if (llvm::cast<IntegerAttr>(attr).getInt() == 1)
         unusedDims.set(dim.index());
 
   // Early exit for the case where the number of unused dims matches the number
   // of ranks reduced.
   if (static_cast<int64_t>(unusedDims.count()) + reducedType.getRank() ==
       originalType.getRank())
     return unusedDims;
llvm::SmallBitVector SubViewOp::getDroppedDims() {
   std::optional<llvm::SmallBitVector> unusedDims =
       computeMemRefRankReductionMask(sourceType, resultType, getMixedSizes());
   assert(unusedDims && "unable to find unused dims of subview");
   return *unusedDims;
 }
 
 OpFoldResult DimOp::fold(FoldAdaptor adaptor) {
   // All forms of folding require a known index.
-  auto index = adaptor.getIndex().dyn_cast_or_null<IntegerAttr>();
+  auto index = llvm::dyn_cast_if_present<IntegerAttr>(adaptor.getIndex());
   if (!index)
     return {};
 
   // Folding for unranked types (UnrankedMemRefType) is not supported.
   auto memrefType = llvm::dyn_cast<MemRefType>(getSource().getType());
   if (!memrefType)
     return {};
 

mlir/lib/Dialect/MemRef/Transforms/ComposeSubView.cpp

LogicalResult matchAndRewrite(memref::SubViewOp op,
     }
 
     // Offsets, sizes and strides OpFoldResult for the combined 'SubViewOp'.
     SmallVector<OpFoldResult> offsets, sizes, strides;
 
     // Because we only support input strides of 1, the output stride is also
     // always 1.
     if (llvm::all_of(strides, [](OpFoldResult &valueOrAttr) {
-          Attribute attr = valueOrAttr.dyn_cast<Attribute>();
+          Attribute attr = llvm::dyn_cast_if_present<Attribute>(valueOrAttr);
           return attr && cast<IntegerAttr>(attr).getInt() == 1;
         })) {
       strides = SmallVector<OpFoldResult>(sourceOp.getMixedStrides().size(),
                                           rewriter.getI64IntegerAttr(1));
     } else {
       return failure();
     }
 
for (auto it : llvm::zip(op.getMixedOffsets(), sourceOp.getMixedOffsets(),
       auto opSize = std::get<2>(it);
 
       // We only support static sizes.
       if (opSize.is<Value>()) {
         return failure();
       }
 
       sizes.push_back(opSize);
-      Attribute opOffsetAttr = opOffset.dyn_cast<Attribute>(),
-                sourceOffsetAttr = sourceOffset.dyn_cast<Attribute>();
+      Attribute opOffsetAttr = llvm::dyn_cast_if_present<Attribute>(opOffset),
+                sourceOffsetAttr =
+                    llvm::dyn_cast_if_present<Attribute>(sourceOffset);
 
       if (opOffsetAttr && sourceOffsetAttr) {
         // If both offsets are static we can simply calculate the combined
         // offset statically.
         offsets.push_back(rewriter.getI64IntegerAttr(
             cast<IntegerAttr>(opOffsetAttr).getInt() +
             cast<IntegerAttr>(sourceOffsetAttr).getInt()));
       } else {
         // When either offset is dynamic, we must emit an additional affine
         // transformation to add the two offsets together dynamically.
         AffineExpr expr = rewriter.getAffineConstantExpr(0);
         SmallVector<Value> affineApplyOperands;
         for (auto valueOrAttr : {opOffset, sourceOffset}) {
-          if (auto attr = valueOrAttr.dyn_cast<Attribute>()) {
+          if (auto attr = llvm::dyn_cast_if_present<Attribute>(valueOrAttr)) {
             expr = expr + cast<IntegerAttr>(attr).getInt();
           } else {
             expr =
                 expr + rewriter.getAffineSymbolExpr(affineApplyOperands.size());
             affineApplyOperands.push_back(valueOrAttr.get<Value>());
           }
         }
 

mlir/lib/Dialect/OpenACC/IR/OpenACC.cpp

checkSymOperandList(Operation *op, std::optional<mlir::ArrayAttr> attributes,
   for (auto args : llvm::zip(operands, *attributes)) {
     mlir::Value operand = std::get<0>(args);
 
     if (!set.insert(operand).second)
       return op->emitOpError()
              << operandName << " operand appears more than once";
 
     mlir::Type varType = operand.getType();
-    auto symbolRef = std::get<1>(args).cast<SymbolRefAttr>();
+    auto symbolRef = llvm::cast<SymbolRefAttr>(std::get<1>(args));
     auto decl = SymbolTable::lookupNearestSymbolFrom<Op>(op, symbolRef);
     if (!decl)
       return op->emitOpError()
              << "expected symbol reference " << symbolRef << " to point to a "
              << operandName << " declaration";
 
     if (decl.getType() && decl.getType() != varType)
       return op->emitOpError() << "expected " << operandName << " (" << varType

mlir/lib/Dialect/OpenMP/IR/OpenMPDialect.cpp

if (!map_types.has_value())
     return success();
 
   if (map_operands.size() != map_types->size())
     return failure();
 
   for (const auto &mapTypeOp : *map_types) {
     int64_t mapTypeBits = 0x00;
 
-    if (!mapTypeOp.isa<mlir::IntegerAttr>())
+    if (!llvm::isa<mlir::IntegerAttr>(mapTypeOp))
       return failure();
 
-    mapTypeBits = mapTypeOp.cast<mlir::IntegerAttr>().getInt();
+    mapTypeBits = llvm::cast<mlir::IntegerAttr>(mapTypeOp).getInt();
 
     bool to =
         bitAnd(mapTypeBits, llvm::omp::OpenMPOffloadMappingFlags::OMP_MAP_TO);
     bool from =
         bitAnd(mapTypeBits, llvm::omp::OpenMPOffloadMappingFlags::OMP_MAP_FROM);
     bool del = bitAnd(mapTypeBits,
                       llvm::omp::OpenMPOffloadMappingFlags::OMP_MAP_DELETE);
 

mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp

static FailureOr<BaseMemRefType> computeLoopRegionIterArgBufferType(
   if (*initArgBufferType == yieldedValueBufferType)
     return yieldedValueBufferType;
 
   // If there is a mismatch between the yielded buffer type and the iter_arg
   // buffer type, the buffer type must be promoted to a fully dynamic layout
   // map.
   auto yieldedRanked = cast<MemRefType>(yieldedValueBufferType);
 #ifndef NDEBUG
-  auto iterRanked = initArgBufferType->cast<MemRefType>();
+  auto iterRanked = llvm::cast<MemRefType>(*initArgBufferType);
   assert(llvm::equal(yieldedRanked.getShape(), iterRanked.getShape()) &&
          "expected same shape");
   assert(yieldedRanked.getMemorySpace() == iterRanked.getMemorySpace() &&
          "expected same memory space");
 #endif // NDEBUG
   return getMemRefTypeWithFullyDynamicLayout(
       cast<RankedTensorType>(iterArg.getType()),
       yieldedRanked.getMemorySpace());
LogicalResult bufferize(Operation *op, RewriterBase &rewriter,
     }
 
     // The result types of a WhileOp are the same as the "after" bbArg types.
     SmallVector<Type> argsTypesAfter = llvm::to_vector(
         llvm::map_range(whileOp.getAfterArguments(), [&](BlockArgument bbArg) {
           if (!isa<TensorType>(bbArg.getType()))
             return bbArg.getType();
           // TODO: error handling
-          return bufferization::getBufferType(bbArg, options)->cast<Type>();
+          return llvm::cast<Type>(*bufferization::getBufferType(bbArg, options));
         }));
 
     // Construct a new scf.while op with memref instead of tensor values.
     ValueRange argsRangeBefore(castedInitArgs);
     TypeRange argsTypesBefore(argsRangeBefore);
     auto newWhileOp = rewriter.create<scf::WhileOp>(
         whileOp.getLoc(), argsTypesAfter, castedInitArgs);
 

mlir/lib/Dialect/SCF/Utils/AffineCanonicalizationUtils.cpp

LogicalResult scf::addLoopRangeConstraints(FlatAffineValueConstraints &cstr,
 
   // IntegerPolyhedron does not support semi-affine expressions.
   // Therefore, only constant step values are supported.
   auto stepInt = getConstantIntValue(step);
   if (!stepInt)
     return failure();
 
   unsigned dimIv = cstr.appendDimVar(iv);
-  auto lbv = lb.dyn_cast<Value>();
+  auto lbv = llvm::dyn_cast_if_present<Value>(lb);
   unsigned symLb =
       lbv ? cstr.appendSymbolVar(lbv) : cstr.appendSymbolVar(/*num=*/1);
-  auto ubv = ub.dyn_cast<Value>();
+  auto ubv = llvm::dyn_cast_if_present<Value>(ub);
   unsigned symUb =
       ubv ? cstr.appendSymbolVar(ubv) : cstr.appendSymbolVar(/*num=*/1);
 
   // If loop lower/upper bounds are constant: Add EQ constraint.
   std::optional<int64_t> lbInt = getConstantIntValue(lb);
   std::optional<int64_t> ubInt = getConstantIntValue(ub);
   if (lbInt)
     cstr.addBound(BoundType::EQ, symLb, *lbInt);

mlir/lib/Dialect/SPIRV/IR/SPIRVCanonicalization.cpp

if (getIndices() == insertOp.getIndices())
       return insertOp.getObject();
   }
 
   if (auto constructOp =
           getComposite().getDefiningOp<spirv::CompositeConstructOp>()) {
     auto type = llvm::cast<spirv::CompositeType>(constructOp.getType());
     if (getIndices().size() == 1 &&
         constructOp.getConstituents().size() == type.getNumElements()) {
-      auto i = getIndices().begin()->cast<IntegerAttr>();
+      auto i = llvm::cast<IntegerAttr>(*getIndices().begin());
       return constructOp.getConstituents()[i.getValue().getSExtValue()];
     }
   }
 
   auto indexVector =
       llvm::to_vector<8>(llvm::map_range(getIndices(), [](Attribute attr) {
         return static_cast<unsigned>(llvm::cast<IntegerAttr>(attr).getInt());
       }));

mlir/lib/Dialect/SPIRV/IR/SPIRVOps.cpp

LogicalResult spirv::BitcastOp::verify() {
   return success();
 }
 
 //===----------------------------------------------------------------------===//
 // spirv.ConvertPtrToUOp
 //===----------------------------------------------------------------------===//
 
 LogicalResult spirv::ConvertPtrToUOp::verify() {
-  auto operandType = getPointer().getType().cast<spirv::PointerType>();
-  auto resultType = getResult().getType().cast<spirv::ScalarType>();
+  auto operandType = llvm::cast<spirv::PointerType>(getPointer().getType());
+  auto resultType = llvm::cast<spirv::ScalarType>(getResult().getType());
   if (!resultType || !resultType.isSignlessInteger())
     return emitError("result must be a scalar type of unsigned integer");
   auto spirvModule = (*this)->getParentOfType<spirv::ModuleOp>();
   if (!spirvModule)
     return success();
   auto addressingModel = spirvModule.getAddressingModel();
   if ((addressingModel == spirv::AddressingModel::Logical) ||
       (addressingModel == spirv::AddressingModel::PhysicalStorageBuffer64 &&
        operandType.getStorageClass() !=
            spirv::StorageClass::PhysicalStorageBuffer))
     return emitError("operand must be a physical pointer");
   return success();
 }
 
 //===----------------------------------------------------------------------===//
 // spirv.ConvertUToPtrOp
 //===----------------------------------------------------------------------===//
 
 LogicalResult spirv::ConvertUToPtrOp::verify() {
-  auto operandType = getOperand().getType().cast<spirv::ScalarType>();
-  auto resultType = getResult().getType().cast<spirv::PointerType>();
+  auto operandType = llvm::cast<spirv::ScalarType>(getOperand().getType());
+  auto resultType = llvm::cast<spirv::PointerType>(getResult().getType());
   if (!operandType || !operandType.isSignlessInteger())
     return emitError("result must be a scalar type of unsigned integer");
   auto spirvModule = (*this)->getParentOfType<spirv::ModuleOp>();
   if (!spirvModule)
     return success();
   auto addressingModel = spirvModule.getAddressingModel();
   if ((addressingModel == spirv::AddressingModel::Logical) ||
       (addressingModel == spirv::AddressingModel::PhysicalStorageBuffer64 &&

mlir/lib/Dialect/SPIRV/IR/SPIRVTypes.cpp

return TypeSwitch<Type, Type>(*this)
       .Case<MatrixType>([](MatrixType type) { return type.getColumnType(); })
       .Case<StructType>(
           [index](StructType type) { return type.getElementType(index); })
       .Default(
           [](Type) -> Type { llvm_unreachable("invalid composite type"); });
 }
 
 unsigned CompositeType::getNumElements() const {
-  if (auto arrayType = dyn_cast<ArrayType>())
+  if (auto arrayType = llvm::dyn_cast<ArrayType>(*this))
     return arrayType.getNumElements();
-  if (auto matrixType = dyn_cast<MatrixType>())
+  if (auto matrixType = llvm::dyn_cast<MatrixType>(*this))
     return matrixType.getNumColumns();
-  if (auto structType = dyn_cast<StructType>())
+  if (auto structType = llvm::dyn_cast<StructType>(*this))
     return structType.getNumElements();
-  if (auto vectorType = dyn_cast<VectorType>())
+  if (auto vectorType = llvm::dyn_cast<VectorType>(*this))
     return vectorType.getNumElements();
-  if (isa<CooperativeMatrixNVType>()) {
+  if (llvm::isa<CooperativeMatrixNVType>(*this)) {
     llvm_unreachable(
         "invalid to query number of elements of spirv::CooperativeMatrix type");
   }
-  if (isa<JointMatrixINTELType>()) {
+  if (llvm::isa<JointMatrixINTELType>(*this)) {
     llvm_unreachable(
         "invalid to query number of elements of spirv::JointMatrix type");
   }
-  if (isa<RuntimeArrayType>()) {
+  if (llvm::isa<RuntimeArrayType>(*this)) {
     llvm_unreachable(
         "invalid to query number of elements of spirv::RuntimeArray type");
   }
   llvm_unreachable("invalid composite type");
 }
 
 bool CompositeType::hasCompileTimeKnownNumElements() const {
-  return !isa<CooperativeMatrixNVType, JointMatrixINTELType,
-              RuntimeArrayType>();
+  return !llvm::isa<CooperativeMatrixNVType, JointMatrixINTELType,
+              RuntimeArrayType>(*this);
 }
 
 void CompositeType::getExtensions(
     SPIRVType::ExtensionArrayRefVector &extensions,
     std::optional<StorageClass> storage) {
   TypeSwitch<Type>(*this)
       .Case<ArrayType, CooperativeMatrixNVType, JointMatrixINTELType,
             MatrixType, RuntimeArrayType, StructType>(
TypeSwitch<Type>(*this)
         }
         return llvm::cast<ScalarType>(type.getElementType())
             .getCapabilities(capabilities, storage);
       })
       .Default([](Type) { llvm_unreachable("invalid composite type"); });
 }
 
 std::optional<int64_t> CompositeType::getSizeInBytes() {
-  if (auto arrayType = dyn_cast<ArrayType>())
+  if (auto arrayType = llvm::dyn_cast<ArrayType>(*this))
     return arrayType.getSizeInBytes();
-  if (auto structType = dyn_cast<StructType>())
+  if (auto structType = llvm::dyn_cast<StructType>(*this))
     return structType.getSizeInBytes();
-  if (auto vectorType = dyn_cast<VectorType>()) {
+  if (auto vectorType = llvm::dyn_cast<VectorType>(*this)) {
     std::optional<int64_t> elementSize =
         llvm::cast<ScalarType>(vectorType.getElementType()).getSizeInBytes();
     if (!elementSize)
       return std::nullopt;
     return *elementSize * vectorType.getNumElements();
   }
   return std::nullopt;
 }

 
 #define WIDTH_CASE(type, width)                                                \
   case width: {                                                                \
     static const Capability caps[] = {Capability::type##width};                \
     ArrayRef<Capability> ref(caps, std::size(caps));                           \
     capabilities.push_back(ref);                                               \
   } break
 
-  if (auto intType = dyn_cast<IntegerType>()) {
+  if (auto intType = llvm::dyn_cast<IntegerType>(*this)) {
     switch (bitwidth) {
       WIDTH_CASE(Int, 8);
       WIDTH_CASE(Int, 16);
       WIDTH_CASE(Int, 64);
     case 1:
     case 32:
       break;
     default:
       llvm_unreachable("invalid bitwidth to getCapabilities");
     }
   } else {
-    assert(isa<FloatType>());
+    assert(llvm::isa<FloatType>(*this));
     switch (bitwidth) {
       WIDTH_CASE(Float, 16);
       WIDTH_CASE(Float, 64);
     case 32:
       break;
     default:
       llvm_unreachable("invalid bitwidth to getCapabilities");
     }
bool SPIRVType::classof(Type type) {
   if (llvm::isa<ScalarType>(type))
     return true;
   if (auto vectorType = llvm::dyn_cast<VectorType>(type))
     return CompositeType::isValid(vectorType);
   return false;
 }
 
 bool SPIRVType::isScalarOrVector() {
-  return isIntOrFloat() || isa<VectorType>();
+  return isIntOrFloat() || llvm::isa<VectorType>(*this);
 }
 
 void SPIRVType::getExtensions(SPIRVType::ExtensionArrayRefVector &extensions,
                               std::optional<StorageClass> storage) {
-  if (auto scalarType = dyn_cast<ScalarType>()) {
+  if (auto scalarType = llvm::dyn_cast<ScalarType>(*this)) {
     scalarType.getExtensions(extensions, storage);
-  } else if (auto compositeType = dyn_cast<CompositeType>()) {
+  } else if (auto compositeType = llvm::dyn_cast<CompositeType>(*this)) {
     compositeType.getExtensions(extensions, storage);
-  } else if (auto imageType = dyn_cast<ImageType>()) {
+  } else if (auto imageType = llvm::dyn_cast<ImageType>(*this)) {
     imageType.getExtensions(extensions, storage);
-  } else if (auto sampledImageType = dyn_cast<SampledImageType>()) {
+  } else if (auto sampledImageType = llvm::dyn_cast<SampledImageType>(*this)) {
     sampledImageType.getExtensions(extensions, storage);
-  } else if (auto matrixType = dyn_cast<MatrixType>()) {
+  } else if (auto matrixType = llvm::dyn_cast<MatrixType>(*this)) {
     matrixType.getExtensions(extensions, storage);
-  } else if (auto ptrType = dyn_cast<PointerType>()) {
+  } else if (auto ptrType = llvm::dyn_cast<PointerType>(*this)) {
     ptrType.getExtensions(extensions, storage);
   } else {
     llvm_unreachable("invalid SPIR-V Type to getExtensions");
   }
 }
 
 void SPIRVType::getCapabilities(
     SPIRVType::CapabilityArrayRefVector &capabilities,
     std::optional<StorageClass> storage) {
-  if (auto scalarType = dyn_cast<ScalarType>()) {
+  if (auto scalarType = llvm::dyn_cast<ScalarType>(*this)) {
     scalarType.getCapabilities(capabilities, storage);
-  } else if (auto compositeType = dyn_cast<CompositeType>()) {
+  } else if (auto compositeType = llvm::dyn_cast<CompositeType>(*this)) {
     compositeType.getCapabilities(capabilities, storage);
-  } else if (auto imageType = dyn_cast<ImageType>()) {
+  } else if (auto imageType = llvm::dyn_cast<ImageType>(*this)) {
     imageType.getCapabilities(capabilities, storage);
-  } else if (auto sampledImageType = dyn_cast<SampledImageType>()) {
+  } else if (auto sampledImageType = llvm::dyn_cast<SampledImageType>(*this)) {
     sampledImageType.getCapabilities(capabilities, storage);
-  } else if (auto matrixType = dyn_cast<MatrixType>()) {
+  } else if (auto matrixType = llvm::dyn_cast<MatrixType>(*this)) {
     matrixType.getCapabilities(capabilities, storage);
-  } else if (auto ptrType = dyn_cast<PointerType>()) {
+  } else if (auto ptrType = llvm::dyn_cast<PointerType>(*this)) {
     ptrType.getCapabilities(capabilities, storage);
   } else {
     llvm_unreachable("invalid SPIR-V Type to getCapabilities");
   }
 }
 
 std::optional<int64_t> SPIRVType::getSizeInBytes() {
-  if (auto scalarType = dyn_cast<ScalarType>())
+  if (auto scalarType = llvm::dyn_cast<ScalarType>(*this))
     return scalarType.getSizeInBytes();
-  if (auto compositeType = dyn_cast<CompositeType>())
+  if (auto compositeType = llvm::dyn_cast<CompositeType>(*this))
     return compositeType.getSizeInBytes();
   return std::nullopt;
 }
 
 //===----------------------------------------------------------------------===//
 // SampledImageType
 //===----------------------------------------------------------------------===//
 struct spirv::detail::SampledImageTypeStorage : public TypeStorage {

mlir/lib/Dialect/Shape/IR/Shape.cpp

 //===----------------------------------------------------------------------===//
 // ConcatOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult ConcatOp::fold(FoldAdaptor adaptor) {
   if (!adaptor.getLhs() || !adaptor.getRhs())
     return nullptr;
   auto lhsShape = llvm::to_vector<6>(
-      adaptor.getLhs().cast<DenseIntElementsAttr>().getValues<int64_t>());
+      llvm::cast<DenseIntElementsAttr>(adaptor.getLhs()).getValues<int64_t>());
   auto rhsShape = llvm::to_vector<6>(
-      adaptor.getRhs().cast<DenseIntElementsAttr>().getValues<int64_t>());
+      llvm::cast<DenseIntElementsAttr>(adaptor.getRhs()).getValues<int64_t>());
   SmallVector<int64_t, 6> resultShape;
   resultShape.append(lhsShape.begin(), lhsShape.end());
   resultShape.append(rhsShape.begin(), rhsShape.end());
   Builder builder(getContext());
   return builder.getIndexTensorAttr(resultShape);
 }
 
 //===----------------------------------------------------------------------===//
if (hasAtMostSingleNonScalar(adaptor.getShapes()))
     return BoolAttr::get(getContext(), true);
 
   if ([&] {
         SmallVector<SmallVector<int64_t, 6>, 6> extents;
         for (const auto &operand : adaptor.getShapes()) {
           if (!operand)
             return false;
           extents.push_back(llvm::to_vector<6>(
-              operand.cast<DenseIntElementsAttr>().getValues<int64_t>()));
+              llvm::cast<DenseIntElementsAttr>(operand).getValues<int64_t>()));
         }
         return OpTrait::util::staticallyKnownBroadcastable(extents);
       }())
     return BoolAttr::get(getContext(), true);
 
   // Lastly, see if folding can be completed based on what constraints are known
   // on the input shapes.
   if ([&] {
LogicalResult mlir::shape::DimOp::verify() {
   return success();
 }
 
 //===----------------------------------------------------------------------===//
 // DivOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult DivOp::fold(FoldAdaptor adaptor) {
-  auto lhs = adaptor.getLhs().dyn_cast_or_null<IntegerAttr>();
+  auto lhs = llvm::dyn_cast_if_present<IntegerAttr>(adaptor.getLhs());
   if (!lhs)
     return nullptr;
-  auto rhs = adaptor.getRhs().dyn_cast_or_null<IntegerAttr>();
+  auto rhs = llvm::dyn_cast_if_present<IntegerAttr>(adaptor.getRhs());
   if (!rhs)
     return nullptr;
 
   // Division in APInt does not follow floor(lhs, rhs) when the result is
   // negative. Rather, APInt rounds toward zero.
   APInt quotient, remainder;
   APInt::sdivrem(lhs.getValue(), rhs.getValue(), quotient, remainder);
   if (quotient.isNegative() && !remainder.isZero()) {
std::optional<int64_t> GetExtentOp::getConstantDim() {
   if (auto constSizeOp = getDim().getDefiningOp<ConstSizeOp>())
     return constSizeOp.getValue().getLimitedValue();
   if (auto constantOp = getDim().getDefiningOp<arith::ConstantOp>())
     return llvm::cast<IntegerAttr>(constantOp.getValue()).getInt();
   return std::nullopt;
 }
 
 OpFoldResult GetExtentOp::fold(FoldAdaptor adaptor) {
-  auto elements = adaptor.getShape().dyn_cast_or_null<DenseIntElementsAttr>();
+  auto elements = llvm::dyn_cast_if_present<DenseIntElementsAttr>(adaptor.getShape());
   if (!elements)
     return nullptr;
   std::optional<int64_t> dim = getConstantDim();
   if (!dim.has_value())
     return nullptr;
   if (dim.value() >= elements.getNumElements())
     return nullptr;
   return elements.getValues<Attribute>()[(uint64_t)dim.value()];
bool mlir::shape::MeetOp::isCompatibleReturnTypes(TypeRange l, TypeRange r) {
   return false;
 }
 
 //===----------------------------------------------------------------------===//
 // RankOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult shape::RankOp::fold(FoldAdaptor adaptor) {
-  auto shape = adaptor.getShape().dyn_cast_or_null<DenseIntElementsAttr>();
+  auto shape = llvm::dyn_cast_if_present<DenseIntElementsAttr>(adaptor.getShape());
   if (!shape)
     return {};
   int64_t rank = shape.getNumElements();
   Builder builder(getContext());
   return builder.getIndexAttr(rank);
 }
 
 /// Evaluate the `rank` operation for shapes of ranked tensors at compile time.
bool mlir::shape::MinOp::isCompatibleReturnTypes(TypeRange l, TypeRange r) {
   return false;
 }
 
 //===----------------------------------------------------------------------===//
 // MulOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult MulOp::fold(FoldAdaptor adaptor) {
-  auto lhs = adaptor.getLhs().dyn_cast_or_null<IntegerAttr>();
+  auto lhs = llvm::dyn_cast_if_present<IntegerAttr>(adaptor.getLhs());
   if (!lhs)
     return nullptr;
-  auto rhs = adaptor.getRhs().dyn_cast_or_null<IntegerAttr>();
+  auto rhs = llvm::dyn_cast_if_present<IntegerAttr>(adaptor.getRhs());
   if (!rhs)
     return nullptr;
   APInt folded = lhs.getValue() * rhs.getValue();
   Type indexTy = IndexType::get(getContext());
   return IntegerAttr::get(indexTy, folded);
 }
 
 LogicalResult mlir::shape::MulOp::inferReturnTypes(

 // SplitAtOp
 //===----------------------------------------------------------------------===//
 
 LogicalResult SplitAtOp::fold(FoldAdaptor adaptor,
                               SmallVectorImpl<OpFoldResult> &results) {
   if (!adaptor.getOperand() || !adaptor.getIndex())
     return failure();
   auto shapeVec = llvm::to_vector<6>(
-      adaptor.getOperand().cast<DenseIntElementsAttr>().getValues<int64_t>());
+      llvm::cast<DenseIntElementsAttr>(adaptor.getOperand()).getValues<int64_t>());
   auto shape = llvm::ArrayRef(shapeVec);
-  auto splitPoint = adaptor.getIndex().cast<IntegerAttr>().getInt();
+  auto splitPoint = llvm::cast<IntegerAttr>(adaptor.getIndex()).getInt();
   // Verify that the split point is in the correct range.
   // TODO: Constant fold to an "error".
   int64_t rank = shape.size();
   if (-rank > splitPoint || splitPoint > rank)
     return failure();
   if (splitPoint < 0)
     splitPoint += shape.size();
   Builder builder(adaptor.getOperand().getContext());
   results.push_back(builder.getIndexTensorAttr(shape.take_front(splitPoint)));
   results.push_back(builder.getIndexTensorAttr(shape.drop_front(splitPoint)));
   return success();
 }
 
 //===----------------------------------------------------------------------===//
 // ToExtentTensorOp
 //===----------------------------------------------------------------------===//
 
 OpFoldResult ToExtentTensorOp::fold(FoldAdaptor adaptor) {
   if (!adaptor.getInput())
     return OpFoldResult();
   Builder builder(getContext());
   auto shape = llvm::to_vector<6>(
-      adaptor.getInput().cast<DenseIntElementsAttr>().getValues<int64_t>());
+      llvm::cast<DenseIntElementsAttr>(adaptor.getInput()).getValues<int64_t>());
   auto type = RankedTensorType::get({static_cast<int64_t>(shape.size())},
                                     builder.getIndexType());
   return DenseIntElementsAttr::get(type, shape);
 }
 
 bool ToExtentTensorOp::areCastCompatible(TypeRange inputs, TypeRange outputs) {
   if (inputs.size() != 1 || outputs.size() != 1)
     return false;

mlir/lib/Dialect/SparseTensor/IR/SparseTensorDialect.cpp

if (!stt.hasEncoding())
     return op->emitError("the sparse-tensor must have an encoding attribute");
   if (!stt.isIdentity())
     return op->emitError("the sparse-tensor must have the identity mapping");
 
   // Verifies the trailing COO.
   Level cooStartLvl = getCOOStart(stt.getEncoding());
   if (cooStartLvl < stt.getLvlRank()) {
     // We only supports trailing COO for now, must be the last input.
-    auto cooTp = lvlTps.back().cast<ShapedType>();
+    auto cooTp = llvm::cast<ShapedType>(lvlTps.back());
     // The coordinates should be in shape of <? x rank>
     unsigned expCOORank = stt.getLvlRank() - cooStartLvl;
     if (cooTp.getRank() != 2 || expCOORank != cooTp.getShape().back()) {
       op->emitError("input/output trailing COO level-ranks don't match");
     }
   }
 
   // Verifies that all types match.
layout.foreachField([&idx, &misMatch, stt, valTp,
     Type inputTp = nullptr;
     if (fKind == SparseTensorFieldKind::ValMemRef) {
       inputTp = valTp;
     } else {
       assert(fid == idx && stt.getLvlType(lvl) == dlt);
       inputTp = lvlTps[idx++];
     }
     // The input element type and expected element type should match.
-    Type inpElemTp = inputTp.cast<TensorType>().getElementType();
+    Type inpElemTp = llvm::cast<TensorType>(inputTp).getElementType();
     Type expElemTp = getFieldElemType(stt, fKind);
     if (inpElemTp != expElemTp) {
       misMatch = true;
       return false; // to terminate the iteration
     }
     return true;
   });
 

mlir/lib/Dialect/SparseTensor/Transforms/SparseGPUCodegen.cpp

static Value genAllocCopy(OpBuilder &builder, Location loc, Value b,
   Value depToken = alloc.getAsyncToken(); // copy-after-alloc
   tokens.push_back(genCopyMemRef(builder, loc, devMem, b, depToken));
   return devMem;
 }
 
 /// Generates a memref from tensor operation.
 static Value genTensorToMemref(PatternRewriter &rewriter, Location loc,
                                Value tensor) {
-  auto tensorType = tensor.getType().cast<ShapedType>();
+  auto tensorType = llvm::cast<ShapedType>(tensor.getType());
   auto memrefType =
       MemRefType::get(tensorType.getShape(), tensorType.getElementType());
   return rewriter.create<bufferization::ToMemrefOp>(loc, memrefType, tensor);
 }
 
 /// Prepares the outlined arguments, passing scalars and buffers in. Here we
 /// assume that the first buffer is the one allocated for output. We create
 /// a set of properly chained asynchronous allocation/copy pairs to increase

mlir/lib/Dialect/SparseTensor/Transforms/SparseTensorCodegen.cpp

public:
   /// insertions occur in "a reasonable order" that enables building the
   /// storage scheme in an appending/inserting kind of fashion (i.e. no
   /// in-between insertions that need data movement). The implementation
   /// relies on CSE/DCE to clean up all bookkeeping that is not needed.
   ///
   /// TODO: better unord/not-unique; also generalize, optimize, specialize!
   SmallVector<Value> genImplementation(TypeRange retTypes, ValueRange args,
                                        OpBuilder &builder, Location loc) {
-    const SparseTensorType stt(rtp.cast<RankedTensorType>());
+    const SparseTensorType stt(llvm::cast<RankedTensorType>(rtp));
     const Level lvlRank = stt.getLvlRank();
     // Extract fields and coordinates from args.
     SmallVector<Value> fields = llvm::to_vector(args.drop_back(lvlRank + 1));
     MutSparseTensorDescriptor desc(stt, fields);
     const SmallVector<Value> coords =
         llvm::to_vector(args.take_back(lvlRank + 1).drop_back());
     Value value = args.back();
     Value parentPos = constantZero(builder, loc, builder.getIndexType());
else
       genStore(builder, loc, value, desc.getValMemRef(), parentPos);
     return fields;
   }
 
   std::string getMangledFuncName() {
     // The mangled name of the function has this format:
     //   <namePrefix>_<DLT>_<shape>_<ordering>_<eltType>_<crdWidth>_<posWidth>
     constexpr const char kInsertFuncNamePrefix[] = "_insert_";
-    const SparseTensorType stt(rtp.cast<RankedTensorType>());
+    const SparseTensorType stt(llvm::cast<RankedTensorType>(rtp));
 
     SmallString<32> nameBuffer;
     llvm::raw_svector_ostream nameOstream(nameBuffer);
     nameOstream << kInsertFuncNamePrefix;
     const Level lvlRank = stt.getLvlRank();
     for (Level l = 0; l < lvlRank; l++)
       nameOstream << toMLIRString(stt.getLvlType(l)) << "_";
     // Static dim sizes are used in the generated code while dynamic sizes are
for (Level l = 0; l < lvlRank; l++) {
     } else {
       assert(isDenseDLT(dlt) || isSingletonDLT(dlt));
     }
   }
 }
 
 static TypedValue<BaseMemRefType> genToMemref(OpBuilder &builder, Location loc,
                                               Value tensor) {
-  auto tTp = tensor.getType().cast<TensorType>();
+  auto tTp = llvm::cast<TensorType>(tensor.getType());
   auto mTp = MemRefType::get(tTp.getShape(), tTp.getElementType());
   return builder.create<bufferization::ToMemrefOp>(loc, mTp, tensor)
       .getResult();
 }
 
 Value genSliceToSize(OpBuilder &builder, Location loc, Value mem, Value sz) {
-  auto elemTp = mem.getType().cast<MemRefType>().getElementType();
+  auto elemTp = llvm::cast<MemRefType>(mem.getType()).getElementType();
   return builder
       .create<memref::SubViewOp>(
           loc, MemRefType::get({ShapedType::kDynamic}, elemTp), mem,
           ValueRange{}, ValueRange{sz}, ValueRange{},
           ArrayRef<int64_t>{0},                    // static offset
           ArrayRef<int64_t>{ShapedType::kDynamic}, // dynamic size
           ArrayRef<int64_t>{1})                    // static stride
       .getResult();

mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp

for (auto dim : llvm::seq<int64_t>(0, padOp.getSourceType().getRank())) {
 
       // Shape along each dimension is source dim + low pad + high pad.
       SmallVector<Value> mapOperands;
       mapOperands.push_back(
           b.createOrFold<tensor::DimOp>(loc, padOp.getSource(), dim));
       AffineExpr expr = b.getAffineDimExpr(0);
       unsigned numSymbols = 0;
       auto addOpFoldResult = [&](OpFoldResult valueOrAttr) {
-        if (Value v = valueOrAttr.dyn_cast<Value>()) {
+        if (Value v = llvm::dyn_cast_if_present<Value>(valueOrAttr)) {
           expr = expr + b.getAffineSymbolExpr(numSymbols++);
           mapOperands.push_back(v);
           return;
         }
         int64_t staticValue =
             llvm::cast<IntegerAttr>(valueOrAttr.get<Attribute>()).getInt();
         expr = expr + staticValue;
       };

mlir/lib/Dialect/Tensor/IR/TensorOps.cpp

Speculation::Speculatability DimOp::getSpeculatability() {
 
   // The verifier rejects operations that violate this assertion.
   assert(constantIndex < rankedSourceType.getRank());
   return Speculation::Speculatable;
 }
 
 OpFoldResult DimOp::fold(FoldAdaptor adaptor) {
   // All forms of folding require a known index.
-  auto index = adaptor.getIndex().dyn_cast_or_null<IntegerAttr>();
+  auto index = llvm::dyn_cast_if_present<IntegerAttr>(adaptor.getIndex());
   if (!index)
     return {};
 
   // Folding for unranked types (UnrankedTensorType) is not supported.
   auto tensorType = llvm::dyn_cast<RankedTensorType>(getSource().getType());
   if (!tensorType)
     return {};
 
LogicalResult matchAndRewrite(CastOp castOp,
     newMixedSizes.reserve(currMixedSizes.size());
     assert(resultShape.size() == currMixedSizes.size() &&
            "mismatch in result shape and sizes of empty op");
     for (auto it : llvm::zip(resultShape, currMixedSizes)) {
       int64_t newDim = std::get<0>(it);
       OpFoldResult currDim = std::get<1>(it);
       // Case 1: The empty tensor dim is static. Check that the tensor cast
       // result dim matches.
-      if (auto attr = currDim.dyn_cast<Attribute>()) {
+      if (auto attr = llvm::dyn_cast_if_present<Attribute>(currDim)) {
         if (ShapedType::isDynamic(newDim) ||
             newDim != llvm::cast<IntegerAttr>(attr).getInt()) {
           // Something is off, the cast result shape cannot be more dynamic
           // than the empty tensor result shape (enforced by
           // `canFoldIntoProducer`). Abort for now.
           return rewriter.notifyMatchFailure(
               producer, "mismatch in static value of shape of empty tensor "
                         "result and cast result");
static Value foldExtractAfterInsertSlice(ExtractSliceOp extractOp) {
   if (insertOp && insertOp.getSource().getType() == extractOp.getType() &&
       insertOp.isSameAs(extractOp, isSame))
     return insertOp.getSource();
 
   return {};
 }
 
 OpFoldResult ExtractSliceOp::fold(FoldAdaptor adaptor) {
-  if (auto splat = adaptor.getSource().dyn_cast_or_null<SplatElementsAttr>()) {
+  if (auto splat = llvm::dyn_cast_if_present<SplatElementsAttr>(adaptor.getSource())) {
     auto resultType = llvm::cast<ShapedType>(getResult().getType());
     if (resultType.hasStaticShape())
       return splat.resizeSplat(resultType);
   }
   if (getSourceType() == getType() &&
       succeeded(foldIdentityOffsetSizeAndStrideOpInterface(*this, getType())))
     return this->getSource();
   if (Value slice = foldExtractAfterInsertSlice(*this))

 
 /// Converts OpFoldResults to int64_t shape entries, unconditionally mapping all
 /// Value's to kDynamic, even if they are arith.constant values.
 static SmallVector<int64_t>
 asShapeWithAnyValueAsDynamic(ArrayRef<OpFoldResult> ofrs) {
   SmallVector<int64_t> result;
   for (auto o : ofrs) {
     // Have to do this first, as getConstantIntValue special-cases constants.
-    if (o.dyn_cast<Value>())
+    if (llvm::dyn_cast_if_present<Value>(o))
       result.push_back(ShapedType::kDynamic);
     else
       result.push_back(getConstantIntValue(o).value_or(ShapedType::kDynamic));
   }
   return result;
 }
 
 /// Helper for PackOp::{getResultShape,inferPackedType}. Returns the shape of

mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp

if (isa<UnrankedTensorType>(castOp.getType())) {
       return getMemRefTypeWithFullyDynamicLayout(castOp.getType(), memorySpace);
     }
 
     // Case 3: Ranked tensor -> ranked tensor. The offsets and strides do not
     // change.
     auto rankedResultType = cast<RankedTensorType>(castOp.getType());
     return MemRefType::get(
         rankedResultType.getShape(), rankedResultType.getElementType(),
-        maybeSrcBufferType->cast<MemRefType>().getLayout(), memorySpace);
+        llvm::cast<MemRefType>(*maybeSrcBufferType).getLayout(), memorySpace);
   }
 
   LogicalResult bufferize(Operation *op, RewriterBase &rewriter,
                           const BufferizationOptions &options) const {
     auto castOp = cast<tensor::CastOp>(op);
 
     // The result buffer still has the old (pre-cast) type.
     FailureOr<Value> resultBuffer =
struct CollapseShapeOpInterface
   FailureOr<BaseMemRefType>
   getBufferType(Operation *op, Value value, const BufferizationOptions &options,
                 const DenseMap<Value, BaseMemRefType> &fixedTypes) const {
     auto collapseShapeOp = cast<tensor::CollapseShapeOp>(op);
     auto maybeSrcBufferType = bufferization::getBufferType(
         collapseShapeOp.getSrc(), options, fixedTypes);
     if (failed(maybeSrcBufferType))
       return failure();
-    auto srcBufferType = maybeSrcBufferType->cast<MemRefType>();
+    auto srcBufferType = llvm::cast<MemRefType>(*maybeSrcBufferType);
     bool canBeCollapsed = memref::CollapseShapeOp::isGuaranteedCollapsible(
         srcBufferType, collapseShapeOp.getReassociationIndices());
 
     if (!canBeCollapsed) {
       // If dims cannot be collapsed, this op bufferizes to a new allocation.
       RankedTensorType tensorResultType = collapseShapeOp.getResultType();
       return bufferization::getMemRefTypeWithStaticIdentityLayout(
           tensorResultType, srcBufferType.getMemorySpace());
struct ExpandShapeOpInterface
   FailureOr<BaseMemRefType>
   getBufferType(Operation *op, Value value, const BufferizationOptions &options,
                 const DenseMap<Value, BaseMemRefType> &fixedTypes) const {
     auto expandShapeOp = cast<tensor::ExpandShapeOp>(op);
     auto maybeSrcBufferType = bufferization::getBufferType(
         expandShapeOp.getSrc(), options, fixedTypes);
     if (failed(maybeSrcBufferType))
       return failure();
-    auto srcBufferType = maybeSrcBufferType->cast<MemRefType>();
+    auto srcBufferType = llvm::cast<MemRefType>(*maybeSrcBufferType);
     auto maybeResultType = memref::ExpandShapeOp::computeExpandedType(
         srcBufferType, expandShapeOp.getResultType().getShape(),
         expandShapeOp.getReassociationIndices());
     if (failed(maybeResultType))
       return failure();
     return *maybeResultType;
   }
 
if (failed(srcMemref))
       return failure();
 
     // Take a subview of the source buffer.
     auto resultMemrefType =
         bufferization::getBufferType(extractSliceOp.getResult(), options);
     if (failed(resultMemrefType))
       return failure();
     Value subView = rewriter.create<memref::SubViewOp>(
-        loc, resultMemrefType->cast<MemRefType>(), *srcMemref, mixedOffsets,
+        loc, llvm::cast<MemRefType>(*resultMemrefType), *srcMemref, mixedOffsets,
         mixedSizes, mixedStrides);
 
     replaceOpWithBufferizedValues(rewriter, op, subView);
     return success();
   }
 
   FailureOr<BaseMemRefType>
   getBufferType(Operation *op, Value value, const BufferizationOptions &options,
                 const DenseMap<Value, BaseMemRefType> &fixedTypes) const {
     auto extractSliceOp = cast<tensor::ExtractSliceOp>(op);
     assert(value == extractSliceOp.getResult() && "invalid value");
     auto srcMemrefType = bufferization::getBufferType(
         extractSliceOp.getSource(), options, fixedTypes);
     if (failed(srcMemrefType))
       return failure();
     SmallVector<OpFoldResult> mixedOffsets = extractSliceOp.getMixedOffsets();
     SmallVector<OpFoldResult> mixedSizes = extractSliceOp.getMixedSizes();
     SmallVector<OpFoldResult> mixedStrides = extractSliceOp.getMixedStrides();
     return cast<BaseMemRefType>(memref::SubViewOp::inferRankReducedResultType(
-        extractSliceOp.getType().getShape(), srcMemrefType->cast<MemRefType>(),
+        extractSliceOp.getType().getShape(), llvm::cast<MemRefType>(*srcMemrefType),
         mixedOffsets, mixedSizes, mixedStrides));
   }
 };
 
 /// Bufferization of tensor.extract. Replace with memref.load.
 struct ExtractOpInterface
     : public BufferizableOpInterface::ExternalModel<ExtractOpInterface,
                                                     tensor::ExtractOp> {

mlir/lib/Dialect/Tosa/IR/TosaCanonicalizations.cpp

 OpFoldResult AddOp::fold(FoldAdaptor adaptor) {
   auto lhsTy = llvm::dyn_cast<RankedTensorType>(getInput1().getType());
   auto rhsTy = llvm::dyn_cast<RankedTensorType>(getInput2().getType());
   auto resultTy = llvm::dyn_cast<RankedTensorType>(getType());
   if (!lhsTy || !rhsTy || !resultTy)
     return {};
 
   auto resultETy = resultTy.getElementType();
-  auto lhsAttr = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
-  auto rhsAttr = adaptor.getInput2().dyn_cast_or_null<DenseElementsAttr>();
+  auto lhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
+  auto rhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput2());
 
   if (lhsTy == resultTy && isSplatZero(resultETy, rhsAttr))
     return getInput1();
   if (rhsTy == resultTy && isSplatZero(resultETy, lhsAttr))
     return getInput2();
 
   if (!lhsAttr || !rhsAttr)
     return {};
 
   return binaryFolder<std::plus<APInt>, std::plus<APFloat>>(lhsAttr, rhsAttr,
                                                             resultTy);
 }
 
 OpFoldResult DivOp::fold(FoldAdaptor adaptor) {
   auto lhsTy = llvm::dyn_cast<RankedTensorType>(getInput1().getType());
   auto rhsTy = llvm::dyn_cast<RankedTensorType>(getInput2().getType());
   auto resultTy = llvm::dyn_cast<RankedTensorType>(getType());
   if (!lhsTy || !rhsTy || !resultTy)
     return {};
   if (lhsTy != rhsTy)
     return {};
 
   auto resultETy = resultTy.getElementType();
-  auto lhsAttr = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
-  auto rhsAttr = adaptor.getInput2().dyn_cast_or_null<DenseElementsAttr>();
+  auto lhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
+  auto rhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput2());
   if (lhsAttr && lhsAttr.isSplat()) {
     if (llvm::isa<IntegerType>(resultETy) &&
         lhsAttr.getSplatValue<APInt>().isZero())
       return lhsAttr;
   }
 
   if (rhsAttr && rhsAttr.isSplat()) {
     if (llvm::isa<IntegerType>(resultETy) &&
OpFoldResult MulOp::fold(FoldAdaptor adaptor) {
   auto rhs = getInput2();
   auto lhsTy = llvm::dyn_cast<RankedTensorType>(lhs.getType());
   auto rhsTy = llvm::dyn_cast<RankedTensorType>(rhs.getType());
   auto resultTy = llvm::dyn_cast<RankedTensorType>(getType());
   if (!lhsTy || !rhsTy || !resultTy)
     return {};
 
   auto resultETy = resultTy.getElementType();
-  auto lhsAttr = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
-  auto rhsAttr = adaptor.getInput2().dyn_cast_or_null<DenseElementsAttr>();
+  auto lhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
+  auto rhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput2());
 
   const int64_t shift = llvm::isa<IntegerType>(resultETy) ? getShift() : 0;
   if (rhsTy == resultTy) {
     if (isSplatZero(resultETy, lhsAttr))
       return lhsAttr.resizeSplat(resultTy);
     if (isSplatOne(resultETy, lhsAttr, shift))
       return rhs;
   }

 OpFoldResult SubOp::fold(FoldAdaptor adaptor) {
   auto lhsTy = llvm::dyn_cast<RankedTensorType>(getInput1().getType());
   auto rhsTy = llvm::dyn_cast<RankedTensorType>(getInput2().getType());
   auto resultTy = llvm::dyn_cast<RankedTensorType>(getType());
   if (!lhsTy || !rhsTy || !resultTy)
     return {};
 
   auto resultETy = resultTy.getElementType();
-  auto lhsAttr = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
-  auto rhsAttr = adaptor.getInput2().dyn_cast_or_null<DenseElementsAttr>();
+  auto lhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
+  auto rhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput2());
 
   if (lhsTy == resultTy && isSplatZero(resultETy, rhsAttr))
     return getInput1();
 
   if (!lhsAttr || !rhsAttr)
     return {};
 
   return binaryFolder<std::minus<APInt>, std::minus<APFloat>>(lhsAttr, rhsAttr,
struct APIntFoldGreaterEqual {
   APInt operator()(const APInt &l, const APInt &r) {
     return APInt(1, l.sge(r));
   }
 };
 } // namespace
 
 OpFoldResult GreaterOp::fold(FoldAdaptor adaptor) {
   auto resultTy = llvm::dyn_cast<RankedTensorType>(getType());
-  auto lhsAttr = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
-  auto rhsAttr = adaptor.getInput2().dyn_cast_or_null<DenseElementsAttr>();
+  auto lhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
+  auto rhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput2());
 
   if (!lhsAttr || !rhsAttr)
     return {};
 
   return binaryFolder<APIntFoldGreater, ComparisonFold<std::greater<APFloat>>>(
       lhsAttr, rhsAttr, resultTy);
 }
 
 OpFoldResult GreaterEqualOp::fold(FoldAdaptor adaptor) {
   auto resultTy = llvm::dyn_cast<RankedTensorType>(getType());
-  auto lhsAttr = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
-  auto rhsAttr = adaptor.getInput2().dyn_cast_or_null<DenseElementsAttr>();
+  auto lhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
+  auto rhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput2());
 
   if (!lhsAttr || !rhsAttr)
     return {};
 
   return binaryFolder<APIntFoldGreaterEqual,
                       ComparisonFold<std::greater_equal<APFloat>>>(
       lhsAttr, rhsAttr, resultTy);
 }
 
 OpFoldResult EqualOp::fold(FoldAdaptor adaptor) {
   auto resultTy = llvm::dyn_cast<RankedTensorType>(getType());
-  auto lhsAttr = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
-  auto rhsAttr = adaptor.getInput2().dyn_cast_or_null<DenseElementsAttr>();
+  auto lhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
+  auto rhsAttr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput2());
   Value lhs = getInput1();
   Value rhs = getInput2();
   auto lhsTy = llvm::cast<ShapedType>(lhs.getType());
 
   // If we are comparing an integer value to itself it is always true. We can
   // not do this with float due to float values.
   if (llvm::isa<IntegerType>(lhsTy.getElementType()) && resultTy &&
       resultTy.hasStaticShape() && lhs == rhs) {
     return DenseElementsAttr::get(resultTy, true);
   }
 
   if (!lhsAttr || !rhsAttr)
     return {};
 
   return binaryFolder<ComparisonFold<std::equal_to<APInt>>,
                       ComparisonFold<std::equal_to<APFloat>>>(lhsAttr, rhsAttr,
                                                               resultTy);
 }
 
 OpFoldResult CastOp::fold(FoldAdaptor adaptor) {
   if (getInput().getType() == getType())
     return getInput();
 
-  auto operand = adaptor.getInput().dyn_cast_or_null<ElementsAttr>();
+  auto operand = llvm::dyn_cast_if_present<ElementsAttr>(adaptor.getInput());
   if (!operand)
     return {};
 
   auto inTy = llvm::cast<ShapedType>(getInput().getType());
   auto outTy = llvm::cast<ShapedType>(getType());
   auto inETy = inTy.getElementType();
   auto outETy = outTy.getElementType();
 
OpFoldResult ReshapeOp::fold(FoldAdaptor adaptor) {
   auto outputTy = llvm::dyn_cast<RankedTensorType>(getType());
 
   if (!inputTy || !outputTy)
     return {};
 
   if (inputTy == outputTy)
     return getInput1();
 
-  auto operand = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>();
+  auto operand = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1());
   if (operand && outputTy.hasStaticShape() && operand.isSplat()) {
     return SplatElementsAttr::get(outputTy, operand.getSplatValue<Attribute>());
   }
 
   return {};
 }
 
 OpFoldResult PadOp::fold(FoldAdaptor adaptor) {
   // If the pad is all zeros we can fold this operation away.
   if (adaptor.getPadding()) {
-    auto densePad = adaptor.getPadding().cast<DenseElementsAttr>();
+    auto densePad = llvm::cast<DenseElementsAttr>(adaptor.getPadding());
     if (densePad.isSplat() && densePad.getSplatValue<APInt>().isZero()) {
       return getInput1();
     }
   }
 
   return {};
 }
 
OpFoldResult ResizeOp::fold(FoldAdaptor adaptor) {
 
   return input;
 }
 
 OpFoldResult ReverseOp::fold(FoldAdaptor adaptor) {
   auto operand = getInput();
   auto operandTy = llvm::cast<ShapedType>(operand.getType());
   auto axis = getAxis();
-  auto operandAttr = adaptor.getInput().dyn_cast_or_null<SplatElementsAttr>();
+  auto operandAttr = llvm::dyn_cast_if_present<SplatElementsAttr>(adaptor.getInput());
   if (operandAttr)
     return operandAttr;
 
   // If the dim-length is 1, tosa.reverse is a no-op.
   if (operandTy.hasRank() && operandTy.getDimSize(axis) == 1)
     return operand;
 
   return {};
OpFoldResult SliceOp::fold(FoldAdaptor adaptor) {
   if (!adaptor.getInput())
     return {};
 
   // Cannot create an ElementsAttr from non-int/float/index types
   if (!inputTy.getElementType().isIntOrIndexOrFloat() ||
       !outputTy.getElementType().isIntOrIndexOrFloat())
     return {};
 
-  auto operand = adaptor.getInput().cast<ElementsAttr>();
+  auto operand = llvm::cast<ElementsAttr>(adaptor.getInput());
   if (operand.isSplat() && outputTy.hasStaticShape()) {
     return SplatElementsAttr::get(outputTy, operand.getSplatValue<Attribute>());
   }
 
   if (inputTy.hasStaticShape() && outputTy.hasStaticShape() &&
       outputTy.getNumElements() == 1) {
     llvm::SmallVector<uint64_t> indices(getStart());
     auto value = operand.getValues<Attribute>()[indices];
     return SplatElementsAttr::get(outputTy, value);
   }
 
   return {};
 }
 
 OpFoldResult tosa::SelectOp::fold(FoldAdaptor adaptor) {
   if (getOnTrue() == getOnFalse())
     return getOnTrue();
 
-  auto predicate = adaptor.getPred().dyn_cast_or_null<DenseIntElementsAttr>();
+  auto predicate = llvm::dyn_cast_if_present<DenseIntElementsAttr>(adaptor.getPred());
   if (!predicate)
     return {};
 
   if (!predicate.isSplat())
     return {};
   return predicate.getSplatValue<APInt>().getBoolValue() ? getOnTrue()
                                                          : getOnFalse();
 }
 
 OpFoldResult TileOp::fold(FoldAdaptor adaptor) {
   bool allOnes = llvm::all_of(getMultiples(), [](int64_t v) { return v == 1; });
   if (allOnes && getInput1().getType() == getType())
     return getInput1();
   return {};
 }
 
 OpFoldResult TransposeOp::fold(FoldAdaptor adaptor) {
   auto inputTy = llvm::cast<ShapedType>(getInput1().getType());
   auto resultTy = llvm::cast<ShapedType>(getType());
 
   // Transposing splat values just means reshaping.
-  if (auto input = adaptor.getInput1().dyn_cast_or_null<DenseElementsAttr>()) {
+  if (auto input = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getInput1())) {
     if (input.isSplat() && resultTy.hasStaticShape() &&
         inputTy.getElementType() == resultTy.getElementType())
       return input.reshape(resultTy);
   }
 
   // Transpose does not change the input type.
   if (getInput1().getType() != getType())
     return {};

mlir/lib/Dialect/Tosa/Transforms/TosaMakeBroadcastable.cpp

LogicalResult reshapeLowerToHigher(PatternRewriter &rewriter, Location loc,
   Value input2_copy = input2;
   if (EqualizeRanks(rewriter, loc, input1_copy, input2_copy).failed()) {
     return rewriter.notifyMatchFailure(loc, "failed to reshape inputs");
   }
 
   // Verify the rank agrees with the output type if the output type is ranked.
   if (outputType) {
     if (outputType.getRank() !=
-            input1_copy.getType().cast<RankedTensorType>().getRank() ||
+            llvm::cast<RankedTensorType>(input1_copy.getType()).getRank() ||
         outputType.getRank() !=
-            input2_copy.getType().cast<RankedTensorType>().getRank())
+            llvm::cast<RankedTensorType>(input2_copy.getType()).getRank())
       return rewriter.notifyMatchFailure(
           loc, "the reshaped type doesn't agrees with the ranked output type");
   }
 
   input1 = input1_copy;
   input2 = input2_copy;
 
   return success();

mlir/lib/Dialect/Tosa/Utils/ConversionUtils.cpp

else if (higherRankDim != lowerRankDim)
       return failure();
   }
   return success();
 }
 } // namespace
 
 LogicalResult mlir::tosa::EqualizeRanks(PatternRewriter &rewriter, Location loc,
                                         Value &input1, Value &input2) {
-  auto input1Ty = input1.getType().dyn_cast<RankedTensorType>();
-  auto input2Ty = input2.getType().dyn_cast<RankedTensorType>();
+  auto input1Ty = llvm::dyn_cast<RankedTensorType>(input1.getType());
+  auto input2Ty = llvm::dyn_cast<RankedTensorType>(input2.getType());
 
   if (!input1Ty || !input2Ty) {
     return failure();
   }
 
   int64_t input1Rank = input1Ty.getRank();
   int64_t input2Rank = input2Ty.getRank();
 
   if (input1Rank == input2Rank)
     return success();
 
   Value higherTensorValue, lowerTensorValue;
   if (input1Rank > input2Rank) {
     higherTensorValue = input1;
     lowerTensorValue = input2;
   } else {
     higherTensorValue = input2;
     lowerTensorValue = input1;
   }
 
   ArrayRef<int64_t> higherRankShape =
-      higherTensorValue.getType().cast<RankedTensorType>().getShape();
+      llvm::cast<RankedTensorType>(higherTensorValue.getType()).getShape();
   ArrayRef<int64_t> lowerRankShape =
-      lowerTensorValue.getType().cast<RankedTensorType>().getShape();
+      llvm::cast<RankedTensorType>(lowerTensorValue.getType()).getShape();
 
   SmallVector<int64_t, 4> reshapeOutputShape;
 
   if (computeReshapeOutput(higherRankShape, lowerRankShape, reshapeOutputShape)
           .failed())
     return failure();
 
-  auto reshapeInputType = lowerTensorValue.getType().cast<RankedTensorType>();
+  auto reshapeInputType =
+      llvm::cast<RankedTensorType>(lowerTensorValue.getType());
   auto reshapeOutputType = RankedTensorType::get(
       ArrayRef<int64_t>(reshapeOutputShape), reshapeInputType.getElementType());
 
   auto reshapeLower = rewriter.create<tosa::ReshapeOp>(
       loc, reshapeOutputType, lowerTensorValue,
       rewriter.getDenseI64ArrayAttr(reshapeOutputShape));
 
   if (input1Rank > input2Rank) {

mlir/lib/Dialect/Transform/IR/TransformInterfaces.cpp

static DiagnosedSilenceableFailure dispatchMappedValues(
     Value handle, ArrayRef<transform::MappedValue> values,
     function_ref<LogicalResult(ArrayRef<Operation *>)> operationsFn,
     function_ref<LogicalResult(ArrayRef<transform::Param>)> paramsFn,
     function_ref<LogicalResult(ValueRange)> valuesFn) {
   if (llvm::isa<transform::TransformHandleTypeInterface>(handle.getType())) {
     SmallVector<Operation *> operations;
     operations.reserve(values.size());
     for (transform::MappedValue value : values) {
-      if (auto *op = value.dyn_cast<Operation *>()) {
+      if (auto *op = llvm::dyn_cast_if_present<Operation *>(value)) {
         operations.push_back(op);
         continue;
       }
       return emitSilenceableFailure(handle.getLoc())
              << "wrong kind of value provided for top-level operation handle";
     }
     if (failed(operationsFn(operations)))
       return DiagnosedSilenceableFailure::definiteFailure();
     return DiagnosedSilenceableFailure::success();
   }
 
   if (llvm::isa<transform::TransformValueHandleTypeInterface>(
           handle.getType())) {
     SmallVector<Value> payloadValues;
     payloadValues.reserve(values.size());
     for (transform::MappedValue value : values) {
-      if (auto v = value.dyn_cast<Value>()) {
+      if (auto v = llvm::dyn_cast_if_present<Value>(value)) {
         payloadValues.push_back(v);
         continue;
       }
       return emitSilenceableFailure(handle.getLoc())
              << "wrong kind of value provided for the top-level value handle";
     }
     if (failed(valuesFn(payloadValues)))
       return DiagnosedSilenceableFailure::definiteFailure();
     return DiagnosedSilenceableFailure::success();
   }
 
   assert(llvm::isa<transform::TransformParamTypeInterface>(handle.getType()) &&
          "unsupported kind of block argument");
   SmallVector<transform::Param> parameters;
   parameters.reserve(values.size());
   for (transform::MappedValue value : values) {
-    if (auto attr = value.dyn_cast<Attribute>()) {
+    if (auto attr = llvm::dyn_cast_if_present<Attribute>(value)) {
       parameters.push_back(attr);
       continue;
     }
     return emitSilenceableFailure(handle.getLoc())
            << "wrong kind of value provided for top-level parameter";
   }
   if (failed(paramsFn(parameters)))
     return DiagnosedSilenceableFailure::definiteFailure();

mlir/lib/Dialect/Utils/StaticValueUtils.cpp

 #include "mlir/Support/MathExtras.h"
 #include "llvm/ADT/APSInt.h"
 
 namespace mlir {
 
 bool isZeroIndex(OpFoldResult v) {
   if (!v)
     return false;
-  if (auto attr = v.dyn_cast<Attribute>()) {
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(v)) {
     IntegerAttr intAttr = dyn_cast<IntegerAttr>(attr);
     return intAttr && intAttr.getValue().isZero();
   }
   if (auto cst = v.get<Value>().getDefiningOp<arith::ConstantIndexOp>())
     return cst.value() == 0;
   return false;
 }
 

 ///   a) it is an IntegerAttr
 /// In other cases, the OpFoldResult is dispached to the `dynamicVec`.
 /// In such dynamic cases, a copy of the `sentinel` value is also pushed to
 /// `staticVec`. This is useful to extract mixed static and dynamic entries that
 /// come from an AttrSizedOperandSegments trait.
 void dispatchIndexOpFoldResult(OpFoldResult ofr,
                                SmallVectorImpl<Value> &dynamicVec,
                                SmallVectorImpl<int64_t> &staticVec) {
-  auto v = ofr.dyn_cast<Value>();
+  auto v = llvm::dyn_cast_if_present<Value>(ofr);
   if (!v) {
     APInt apInt = cast<IntegerAttr>(ofr.get<Attribute>()).getValue();
     staticVec.push_back(apInt.getSExtValue());
     return;
   }
   dynamicVec.push_back(v);
   staticVec.push_back(ShapedType::kDynamic);
 }
SmallVector<OpFoldResult> getAsIndexOpFoldResult(MLIRContext *ctx,
                                                  ArrayRef<int64_t> values) {
   return llvm::to_vector(llvm::map_range(
       values, [ctx](int64_t v) { return getAsIndexOpFoldResult(ctx, v); }));
 }
 
 /// If ofr is a constant integer or an IntegerAttr, return the integer.
 std::optional<int64_t> getConstantIntValue(OpFoldResult ofr) {
   // Case 1: Check for Constant integer.
-  if (auto val = ofr.dyn_cast<Value>()) {
+  if (auto val = llvm::dyn_cast_if_present<Value>(ofr)) {
     APSInt intVal;
     if (matchPattern(val, m_ConstantInt(&intVal)))
       return intVal.getSExtValue();
     return std::nullopt;
   }
   // Case 2: Check for IntegerAttr.
-  Attribute attr = ofr.dyn_cast<Attribute>();
+  Attribute attr = llvm::dyn_cast_if_present<Attribute>(ofr);
   if (auto intAttr = dyn_cast_or_null<IntegerAttr>(attr))
     return intAttr.getValue().getSExtValue();
   return std::nullopt;
 }
 
 /// Return true if `ofr` is constant integer equal to `value`.
 bool isConstantIntValue(OpFoldResult ofr, int64_t value) {
   auto val = getConstantIntValue(ofr);
   return val && *val == value;
 }
 
 /// Return true if ofr1 and ofr2 are the same integer constant attribute values
 /// or the same SSA value.
 /// Ignore integer bitwidth and type mismatch that come from the fact there is
 /// no IndexAttr and that IndexType has no bitwidth.
 bool isEqualConstantIntOrValue(OpFoldResult ofr1, OpFoldResult ofr2) {
   auto cst1 = getConstantIntValue(ofr1), cst2 = getConstantIntValue(ofr2);
   if (cst1 && cst2 && *cst1 == *cst2)
     return true;
-  auto v1 = ofr1.dyn_cast<Value>(), v2 = ofr2.dyn_cast<Value>();
+  auto v1 = llvm::dyn_cast_if_present<Value>(ofr1),
+       v2 = llvm::dyn_cast_if_present<Value>(ofr2);
   return v1 && v1 == v2;
 }
 
 bool isEqualConstantIntOrValueArray(ArrayRef<OpFoldResult> ofrs1,
                                     ArrayRef<OpFoldResult> ofrs2) {
   if (ofrs1.size() != ofrs2.size())
     return false;
   for (auto [ofr1, ofr2] : llvm::zip_equal(ofrs1, ofrs2))

mlir/lib/Dialect/Vector/IR/VectorOps.cpp

 }
 
 LogicalResult
 ExtractOp::inferReturnTypes(MLIRContext *, std::optional<Location>,
                             ValueRange operands, DictionaryAttr attributes,
                             OpaqueProperties properties, RegionRange,
                             SmallVectorImpl<Type> &inferredReturnTypes) {
   ExtractOp::Adaptor op(operands, attributes, properties);
-  auto vectorType = op.getVector().getType().cast<VectorType>();
+  auto vectorType = llvm::cast<VectorType>(op.getVector().getType());
   if (static_cast<int64_t>(op.getPosition().size()) == vectorType.getRank()) {
     inferredReturnTypes.push_back(vectorType.getElementType());
   } else {
     auto n =
         std::min<size_t>(op.getPosition().size(), vectorType.getRank() - 1);
     inferredReturnTypes.push_back(VectorType::get(
         vectorType.getShape().drop_front(n), vectorType.getElementType()));
   }

 }
 
 OpFoldResult BroadcastOp::fold(FoldAdaptor adaptor) {
   if (getSourceType() == getResultVectorType())
     return getSource();
   if (!adaptor.getSource())
     return {};
   auto vectorType = getResultVectorType();
-  if (adaptor.getSource().isa<IntegerAttr, FloatAttr>())
+  if (llvm::isa<IntegerAttr, FloatAttr>(adaptor.getSource()))
     return DenseElementsAttr::get(vectorType, adaptor.getSource());
-  if (auto attr = adaptor.getSource().dyn_cast<SplatElementsAttr>())
+  if (auto attr = llvm::dyn_cast<SplatElementsAttr>(adaptor.getSource()))
     return DenseElementsAttr::get(vectorType, attr.getSplatValue<Attribute>());
   return {};
 }
 
 namespace {
 
 // Fold broadcast1(broadcast2(x)) into broadcast1(x).
 struct BroadcastFolder : public OpRewritePattern<BroadcastOp> {

 }
 
 LogicalResult
 ShuffleOp::inferReturnTypes(MLIRContext *, std::optional<Location>,
                             ValueRange operands, DictionaryAttr attributes,
                             OpaqueProperties properties, RegionRange,
                             SmallVectorImpl<Type> &inferredReturnTypes) {
   ShuffleOp::Adaptor op(operands, attributes, properties);
-  auto v1Type = op.getV1().getType().cast<VectorType>();
+  auto v1Type = llvm::cast<VectorType>(op.getV1().getType());
   auto v1Rank = v1Type.getRank();
   // Construct resulting type: leading dimension matches mask
   // length, all trailing dimensions match the operands.
   SmallVector<int64_t, 4> shape;
   shape.reserve(v1Rank);
   shape.push_back(std::max<size_t>(1, op.getMask().size()));
   // In the 0-D case there is no trailing shape to append.
   if (v1Rank > 0)
void vector::TransposeOp::build(OpBuilder &builder, OperationState &result,
 
   result.addOperands(vector);
   result.addTypes(VectorType::get(transposedShape, vt.getElementType()));
   result.addAttribute(getTranspAttrStrName(), builder.getI64ArrayAttr(transp));
 }
 
 OpFoldResult vector::TransposeOp::fold(FoldAdaptor adaptor) {
   // Eliminate splat constant transpose ops.
-  if (auto attr = adaptor.getVector().dyn_cast_or_null<DenseElementsAttr>())
+  if (auto attr = llvm::dyn_cast_if_present<DenseElementsAttr>(adaptor.getVector()))
     if (attr.isSplat())
       return attr.reshape(getResultVectorType());
 
   // Eliminate identity transpose ops. This happens when the dimensions of the
   // input vector remain in their original order after the transpose operation.
   SmallVector<int64_t, 4> transp;
   getTransp(transp);
 

mlir/lib/IR/AsmPrinter.cpp

void Value::print(raw_ostream &os, const OpPrintingFlags &flags) {
   if (!impl) {
     os << "<<NULL VALUE>>";
     return;
   }
 
   if (auto *op = getDefiningOp())
     return op->print(os, flags);
   // TODO: Improve BlockArgument print'ing.
-  BlockArgument arg = this->cast<BlockArgument>();
+  BlockArgument arg = llvm::cast<BlockArgument>(*this);
   os << "<block argument> of type '" << arg.getType()
      << "' at index: " << arg.getArgNumber();
 }
 void Value::print(raw_ostream &os, AsmState &state) {
   if (!impl) {
     os << "<<NULL VALUE>>";
     return;
   }
 
   if (auto *op = getDefiningOp())
     return op->print(os, state);
 
   // TODO: Improve BlockArgument print'ing.
-  BlockArgument arg = this->cast<BlockArgument>();
+  BlockArgument arg = llvm::cast<BlockArgument>(*this);
   os << "<block argument> of type '" << arg.getType()
      << "' at index: " << arg.getArgNumber();
 }
 
 void Value::dump() {
   print(llvm::errs());
   llvm::errs() << "\n";
 }
if (!parentOp)
       break;
     op = parentOp;
   } while (true);
   return op;
 }
 
 void Value::printAsOperand(raw_ostream &os, const OpPrintingFlags &flags) {
   Operation *op;
-  if (auto result = dyn_cast<OpResult>()) {
+  if (auto result = llvm::dyn_cast<OpResult>(*this)) {
     op = result.getOwner();
   } else {
-    op = cast<BlockArgument>().getOwner()->getParentOp();
+    op = llvm::cast<BlockArgument>(*this).getOwner()->getParentOp();
     if (!op) {
       os << "<<UNKNOWN SSA VALUE>>";
       return;
     }
   }
   op = findParent(op, flags.shouldUseLocalScope());
   AsmState state(op, flags);
   printAsOperand(os, state);

mlir/lib/IR/Block.cpp

if ((count = blocks.size()))
     base = blocks.data();
 }
 
 BlockRange::BlockRange(SuccessorRange successors)
     : BlockRange(successors.begin().getBase(), successors.size()) {}
 
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 BlockRange::OwnerT BlockRange::offset_base(OwnerT object, ptrdiff_t index) {
-  if (auto *operand = object.dyn_cast<BlockOperand *>())
+  if (auto *operand = llvm::dyn_cast_if_present<BlockOperand *>(object))
     return {operand + index};
-  return {object.dyn_cast<Block *const *>() + index};
+  return {llvm::dyn_cast_if_present<Block *const *>(object) + index};
 }
 
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 Block *BlockRange::dereference_iterator(OwnerT object, ptrdiff_t index) {
-  if (const auto *operand = object.dyn_cast<BlockOperand *>())
+  if (const auto *operand = llvm::dyn_cast_if_present<BlockOperand *>(object))
     return operand[index].get();
-  return object.dyn_cast<Block *const *>()[index];
+  return llvm::dyn_cast_if_present<Block *const *>(object)[index];
 }

mlir/lib/IR/Builders.cpp

LogicalResult OpBuilder::tryFold(Operation *op,
   SmallVector<Operation *, 1> generatedConstants;
 
   // Populate the results with the folded results.
   Dialect *dialect = op->getDialect();
   for (auto it : llvm::zip(foldResults, opResults.getTypes())) {
     Type expectedType = std::get<1>(it);
 
     // Normal values get pushed back directly.
-    if (auto value = std::get<0>(it).dyn_cast<Value>()) {
+    if (auto value = llvm::dyn_cast_if_present<Value>(std::get<0>(it))) {
       if (value.getType() != expectedType)
         return cleanupFailure();
 
       results.push_back(value);
       continue;
     }
 
     // Otherwise, try to materialize a constant operation.

mlir/lib/IR/BuiltinAttributes.cpp

assert(getDenseElementBitWidth(newElType) ==
          "expected element types with the same bitwidth");
   return DenseIntOrFPElementsAttr::getRaw(curType.clone(newElType),
                                           getRawData());
 }
 
 DenseElementsAttr
 DenseElementsAttr::mapValues(Type newElementType,
                              function_ref<APInt(const APInt &)> mapping) const {
-  return cast<DenseIntElementsAttr>().mapValues(newElementType, mapping);
+  return llvm::cast<DenseIntElementsAttr>(*this).mapValues(newElementType, mapping);
 }
 
 DenseElementsAttr DenseElementsAttr::mapValues(
     Type newElementType, function_ref<APInt(const APFloat &)> mapping) const {
-  return cast<DenseFPElementsAttr>().mapValues(newElementType, mapping);
+  return llvm::cast<DenseFPElementsAttr>(*this).mapValues(newElementType, mapping);
 }
 
 ShapedType DenseElementsAttr::getType() const {
   return static_cast<const DenseElementsAttributeStorage *>(impl)->type;
 }
 
 Type DenseElementsAttr::getElementType() const {
   return getType().getElementType();

mlir/lib/IR/BuiltinTypes.cpp

IntegerType IntegerType::scaleElementBitwidth(unsigned scale) {
   return IntegerType::get(getContext(), scale * getWidth(), getSignedness());
 }
 
 //===----------------------------------------------------------------------===//
 // Float Type
 //===----------------------------------------------------------------------===//
 
 unsigned FloatType::getWidth() {
-  if (isa<Float8E5M2Type, Float8E4M3FNType, Float8E5M2FNUZType,
-          Float8E4M3FNUZType, Float8E4M3B11FNUZType>())
+  if (llvm::isa<Float8E5M2Type, Float8E4M3FNType, Float8E5M2FNUZType,
+          Float8E4M3FNUZType, Float8E4M3B11FNUZType>(*this))
     return 8;
-  if (isa<Float16Type, BFloat16Type>())
+  if (llvm::isa<Float16Type, BFloat16Type>(*this))
     return 16;
-  if (isa<Float32Type>())
+  if (llvm::isa<Float32Type>(*this))
     return 32;
-  if (isa<Float64Type>())
+  if (llvm::isa<Float64Type>(*this))
     return 64;
-  if (isa<Float80Type>())
+  if (llvm::isa<Float80Type>(*this))
     return 80;
-  if (isa<Float128Type>())
+  if (llvm::isa<Float128Type>(*this))
     return 128;
   llvm_unreachable("unexpected float type");
 }
 
 /// Returns the floating semantics for the given type.
 const llvm::fltSemantics &FloatType::getFloatSemantics() {
-  if (isa<Float8E5M2Type>())
+  if (llvm::isa<Float8E5M2Type>(*this))
     return APFloat::Float8E5M2();
-  if (isa<Float8E4M3FNType>())
+  if (llvm::isa<Float8E4M3FNType>(*this))
     return APFloat::Float8E4M3FN();
-  if (isa<Float8E5M2FNUZType>())
+  if (llvm::isa<Float8E5M2FNUZType>(*this))
     return APFloat::Float8E5M2FNUZ();
-  if (isa<Float8E4M3FNUZType>())
+  if (llvm::isa<Float8E4M3FNUZType>(*this))
     return APFloat::Float8E4M3FNUZ();
-  if (isa<Float8E4M3B11FNUZType>())
+  if (llvm::isa<Float8E4M3B11FNUZType>(*this))
     return APFloat::Float8E4M3B11FNUZ();
-  if (isa<BFloat16Type>())
+  if (llvm::isa<BFloat16Type>(*this))
     return APFloat::BFloat();
-  if (isa<Float16Type>())
+  if (llvm::isa<Float16Type>(*this))
     return APFloat::IEEEhalf();
-  if (isa<Float32Type>())
+  if (llvm::isa<Float32Type>(*this))
     return APFloat::IEEEsingle();
-  if (isa<Float64Type>())
+  if (llvm::isa<Float64Type>(*this))
     return APFloat::IEEEdouble();
-  if (isa<Float80Type>())
+  if (llvm::isa<Float80Type>(*this))
     return APFloat::x87DoubleExtended();
-  if (isa<Float128Type>())
+  if (llvm::isa<Float128Type>(*this))
     return APFloat::IEEEquad();
   llvm_unreachable("non-floating point type used");
 }
 
 FloatType FloatType::scaleElementBitwidth(unsigned scale) {
   if (!scale)
     return FloatType();
   MLIRContext *ctx = getContext();

 //===----------------------------------------------------------------------===//
 
 Type TensorType::getElementType() const {
   return llvm::TypeSwitch<TensorType, Type>(*this)
       .Case<RankedTensorType, UnrankedTensorType>(
           [](auto type) { return type.getElementType(); });
 }
 
-bool TensorType::hasRank() const { return !isa<UnrankedTensorType>(); }
+bool TensorType::hasRank() const { return !llvm::isa<UnrankedTensorType>(*this); }
 
 ArrayRef<int64_t> TensorType::getShape() const {
-  return cast<RankedTensorType>().getShape();
+  return llvm::cast<RankedTensorType>(*this).getShape();
 }
 
 TensorType TensorType::cloneWith(std::optional<ArrayRef<int64_t>> shape,
                                  Type elementType) const {
-  if (auto unrankedTy = dyn_cast<UnrankedTensorType>()) {
+  if (auto unrankedTy = llvm::dyn_cast<UnrankedTensorType>(*this)) {
     if (shape)
       return RankedTensorType::get(*shape, elementType);
     return UnrankedTensorType::get(elementType);
   }
 
-  auto rankedTy = cast<RankedTensorType>();
+  auto rankedTy = llvm::cast<RankedTensorType>(*this);
   if (!shape)
     return RankedTensorType::get(rankedTy.getShape(), elementType,
                                  rankedTy.getEncoding());
   return RankedTensorType::get(shape.value_or(rankedTy.getShape()), elementType,
                                rankedTy.getEncoding());
 }
 
 RankedTensorType TensorType::clone(::llvm::ArrayRef<int64_t> shape,

 //===----------------------------------------------------------------------===//
 
 Type BaseMemRefType::getElementType() const {
   return llvm::TypeSwitch<BaseMemRefType, Type>(*this)
       .Case<MemRefType, UnrankedMemRefType>(
           [](auto type) { return type.getElementType(); });
 }
 
-bool BaseMemRefType::hasRank() const { return !isa<UnrankedMemRefType>(); }
+bool BaseMemRefType::hasRank() const { return !llvm::isa<UnrankedMemRefType>(*this); }
 
 ArrayRef<int64_t> BaseMemRefType::getShape() const {
-  return cast<MemRefType>().getShape();
+  return llvm::cast<MemRefType>(*this).getShape();
 }
 
 BaseMemRefType BaseMemRefType::cloneWith(std::optional<ArrayRef<int64_t>> shape,
                                          Type elementType) const {
-  if (auto unrankedTy = dyn_cast<UnrankedMemRefType>()) {
+  if (auto unrankedTy = llvm::dyn_cast<UnrankedMemRefType>(*this)) {
     if (!shape)
       return UnrankedMemRefType::get(elementType, getMemorySpace());
     MemRefType::Builder builder(*shape, elementType);
     builder.setMemorySpace(getMemorySpace());
     return builder;
   }
 
-  MemRefType::Builder builder(cast<MemRefType>());
+  MemRefType::Builder builder(llvm::cast<MemRefType>(*this));
   if (shape)
     builder.setShape(*shape);
   builder.setElementType(elementType);
   return builder;
 }
 
 MemRefType BaseMemRefType::clone(::llvm::ArrayRef<int64_t> shape,
                                  Type elementType) const {
   return ::llvm::cast<MemRefType>(cloneWith(shape, elementType));
 }
 
 MemRefType BaseMemRefType::clone(::llvm::ArrayRef<int64_t> shape) const {
   return ::llvm::cast<MemRefType>(cloneWith(shape, getElementType()));
 }
 
 Attribute BaseMemRefType::getMemorySpace() const {
-  if (auto rankedMemRefTy = dyn_cast<MemRefType>())
+  if (auto rankedMemRefTy = llvm::dyn_cast<MemRefType>(*this))
     return rankedMemRefTy.getMemorySpace();
-  return cast<UnrankedMemRefType>().getMemorySpace();
+  return llvm::cast<UnrankedMemRefType>(*this).getMemorySpace();
 }
 
 unsigned BaseMemRefType::getMemorySpaceAsInt() const {
-  if (auto rankedMemRefTy = dyn_cast<MemRefType>())
+  if (auto rankedMemRefTy = llvm::dyn_cast<MemRefType>(*this))
     return rankedMemRefTy.getMemorySpaceAsInt();
-  return cast<UnrankedMemRefType>().getMemorySpaceAsInt();
+  return llvm::cast<UnrankedMemRefType>(*this).getMemorySpaceAsInt();
 }
 
 //===----------------------------------------------------------------------===//
 // MemRefType
 //===----------------------------------------------------------------------===//
 
 /// Given an `originalShape` and a `reducedShape` assumed to be a subset of
 /// `originalShape` with some `1` entries erased, return the set of indices

mlir/lib/IR/OperationSupport.cpp

 ValueRange::ValueRange(OperandRange values)
     : ValueRange(values.begin().getBase(), values.size()) {}
 ValueRange::ValueRange(ResultRange values)
     : ValueRange(values.getBase(), values.size()) {}
 
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 ValueRange::OwnerT ValueRange::offset_base(const OwnerT &owner,
                                            ptrdiff_t index) {
-  if (const auto *value = owner.dyn_cast<const Value *>())
+  if (const auto *value = llvm::dyn_cast_if_present<const Value *>(owner))
     return {value + index};
-  if (auto *operand = owner.dyn_cast<OpOperand *>())
+  if (auto *operand = llvm::dyn_cast_if_present<OpOperand *>(owner))
     return {operand + index};
   return owner.get<detail::OpResultImpl *>()->getNextResultAtOffset(index);
 }
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 Value ValueRange::dereference_iterator(const OwnerT &owner, ptrdiff_t index) {
-  if (const auto *value = owner.dyn_cast<const Value *>())
+  if (const auto *value = llvm::dyn_cast_if_present<const Value *>(owner))
     return value[index];
-  if (auto *operand = owner.dyn_cast<OpOperand *>())
+  if (auto *operand = llvm::dyn_cast_if_present<OpOperand *>(owner))
     return operand[index].get();
   return owner.get<detail::OpResultImpl *>()->getNextResultAtOffset(index);
 }
 
 //===----------------------------------------------------------------------===//
 // Operation Equivalency
 //===----------------------------------------------------------------------===//
 

mlir/lib/IR/Region.cpp

 RegionRange::RegionRange(ArrayRef<std::unique_ptr<Region>> regions)
     : RegionRange(regions.data(), regions.size()) {}
 RegionRange::RegionRange(ArrayRef<Region *> regions)
     : RegionRange(const_cast<Region **>(regions.data()), regions.size()) {}
 
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 RegionRange::OwnerT RegionRange::offset_base(const OwnerT &owner,
                                              ptrdiff_t index) {
-  if (auto *region = owner.dyn_cast<const std::unique_ptr<Region> *>())
+  if (auto *region = llvm::dyn_cast_if_present<const std::unique_ptr<Region> *>(owner))
     return region + index;
-  if (auto **region = owner.dyn_cast<Region **>())
+  if (auto **region = llvm::dyn_cast_if_present<Region **>(owner))
     return region + index;
   return &owner.get<Region *>()[index];
 }
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 Region *RegionRange::dereference_iterator(const OwnerT &owner,
                                           ptrdiff_t index) {
-  if (auto *region = owner.dyn_cast<const std::unique_ptr<Region> *>())
+  if (auto *region = llvm::dyn_cast_if_present<const std::unique_ptr<Region> *>(owner))
     return region[index].get();
-  if (auto **region = owner.dyn_cast<Region **>())
+  if (auto **region = llvm::dyn_cast_if_present<Region **>(owner))
     return region[index];
   return &owner.get<Region *>()[index];
 }

mlir/lib/IR/SymbolTable.cpp

struct SymbolScope {
   /// Walk the symbol uses within this scope, invoking the given callback.
   /// This variant is used when the callback type matches that expected by
   /// 'walkSymbolUses'.
   template <typename CallbackT,
             std::enable_if_t<!std::is_same<
                 typename llvm::function_traits<CallbackT>::result_t,
                 void>::value> * = nullptr>
   std::optional<WalkResult> walk(CallbackT cback) {
-    if (Region *region = limit.dyn_cast<Region *>())
+    if (Region *region = llvm::dyn_cast_if_present<Region *>(limit))
       return walkSymbolUses(*region, cback);
     return walkSymbolUses(limit.get<Operation *>(), cback);
   }
   /// This variant is used when the callback type matches a stripped down type:
   /// void(SymbolTable::SymbolUse use)
   template <typename CallbackT,
             std::enable_if_t<std::is_same<
                 typename llvm::function_traits<CallbackT>::result_t,
                 void>::value> * = nullptr>
   std::optional<WalkResult> walk(CallbackT cback) {
     return walk([=](SymbolTable::SymbolUse use) {
       return cback(use), WalkResult::advance();
     });
   }
 
   /// Walk all of the operations nested under the current scope without
   /// traversing into any nested symbol tables.
   template <typename CallbackT>
   std::optional<WalkResult> walkSymbolTable(CallbackT &&cback) {
-    if (Region *region = limit.dyn_cast<Region *>())
+    if (Region *region = llvm::dyn_cast_if_present<Region *>(limit))
       return ::walkSymbolTable(*region, cback);
     return ::walkSymbolTable(limit.get<Operation *>(), cback);
   }
 
   /// The representation of the symbol within this scope.
   SymbolRefAttr symbol;
 
   /// The IR unit representing this scope.

mlir/lib/IR/TypeRange.cpp

 TypeRange::TypeRange(OperandRange values)
     : TypeRange(values.begin().getBase(), values.size()) {}
 TypeRange::TypeRange(ResultRange values)
     : TypeRange(values.getBase(), values.size()) {}
 TypeRange::TypeRange(ValueRange values) : TypeRange(OwnerT(), values.size()) {
   if (count == 0)
     return;
   ValueRange::OwnerT owner = values.begin().getBase();
-  if (auto *result = owner.dyn_cast<detail::OpResultImpl *>())
+  if (auto *result = llvm::dyn_cast_if_present<detail::OpResultImpl *>(owner))
     this->base = result;
-  else if (auto *operand = owner.dyn_cast<OpOperand *>())
+  else if (auto *operand = llvm::dyn_cast_if_present<OpOperand *>(owner))
     this->base = operand;
   else
     this->base = owner.get<const Value *>();
 }
 
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 TypeRange::OwnerT TypeRange::offset_base(OwnerT object, ptrdiff_t index) {
-  if (const auto *value = object.dyn_cast<const Value *>())
+  if (const auto *value = llvm::dyn_cast_if_present<const Value *>(object))
     return {value + index};
-  if (auto *operand = object.dyn_cast<OpOperand *>())
+  if (auto *operand = llvm::dyn_cast_if_present<OpOperand *>(object))
     return {operand + index};
-  if (auto *result = object.dyn_cast<detail::OpResultImpl *>())
+  if (auto *result = llvm::dyn_cast_if_present<detail::OpResultImpl *>(object))
     return {result->getNextResultAtOffset(index)};
-  return {object.dyn_cast<const Type *>() + index};
+  return {llvm::dyn_cast_if_present<const Type *>(object) + index};
 }
 
 /// See `llvm::detail::indexed_accessor_range_base` for details.
 Type TypeRange::dereference_iterator(OwnerT object, ptrdiff_t index) {
-  if (const auto *value = object.dyn_cast<const Value *>())
+  if (const auto *value = llvm::dyn_cast_if_present<const Value *>(object))
     return (value + index)->getType();
-  if (auto *operand = object.dyn_cast<OpOperand *>())
+  if (auto *operand = llvm::dyn_cast_if_present<OpOperand *>(object))
     return (operand + index)->get().getType();
-  if (auto *result = object.dyn_cast<detail::OpResultImpl *>())
+  if (auto *result = llvm::dyn_cast_if_present<detail::OpResultImpl *>(object))
     return result->getNextResultAtOffset(index)->getType();
-  return object.dyn_cast<const Type *>()[index];
+  return llvm::dyn_cast_if_present<const Type *>(object)[index];
 }

mlir/lib/IR/Types.cpp

 }
 
 //===----------------------------------------------------------------------===//
 // Type
 //===----------------------------------------------------------------------===//
 
 MLIRContext *Type::getContext() const { return getDialect().getContext(); }
 
-bool Type::isFloat8E5M2() const { return isa<Float8E5M2Type>(); }
-bool Type::isFloat8E4M3FN() const { return isa<Float8E4M3FNType>(); }
-bool Type::isFloat8E5M2FNUZ() const { return isa<Float8E5M2FNUZType>(); }
-bool Type::isFloat8E4M3FNUZ() const { return isa<Float8E4M3FNUZType>(); }
-bool Type::isFloat8E4M3B11FNUZ() const { return isa<Float8E4M3B11FNUZType>(); }
-bool Type::isBF16() const { return isa<BFloat16Type>(); }
-bool Type::isF16() const { return isa<Float16Type>(); }
-bool Type::isF32() const { return isa<Float32Type>(); }
-bool Type::isF64() const { return isa<Float64Type>(); }
-bool Type::isF80() const { return isa<Float80Type>(); }
-bool Type::isF128() const { return isa<Float128Type>(); }
+bool Type::isFloat8E5M2() const { return llvm::isa<Float8E5M2Type>(*this); }
+bool Type::isFloat8E4M3FN() const { return llvm::isa<Float8E4M3FNType>(*this); }
+bool Type::isFloat8E5M2FNUZ() const {
+  return llvm::isa<Float8E5M2FNUZType>(*this);
+}
+bool Type::isFloat8E4M3FNUZ() const {
+  return llvm::isa<Float8E4M3FNUZType>(*this);
+}
+bool Type::isFloat8E4M3B11FNUZ() const {
+  return llvm::isa<Float8E4M3B11FNUZType>(*this);
+}
+bool Type::isBF16() const { return llvm::isa<BFloat16Type>(*this); }
+bool Type::isF16() const { return llvm::isa<Float16Type>(*this); }
+bool Type::isF32() const { return llvm::isa<Float32Type>(*this); }
+bool Type::isF64() const { return llvm::isa<Float64Type>(*this); }
+bool Type::isF80() const { return llvm::isa<Float80Type>(*this); }
+bool Type::isF128() const { return llvm::isa<Float128Type>(*this); }
 
-bool Type::isIndex() const { return isa<IndexType>(); }
+bool Type::isIndex() const { return llvm::isa<IndexType>(*this); }
 
 /// Return true if this is an integer type with the specified width.
 bool Type::isInteger(unsigned width) const {
-  if (auto intTy = dyn_cast<IntegerType>())
+  if (auto intTy = llvm::dyn_cast<IntegerType>(*this))
     return intTy.getWidth() == width;
   return false;
 }
 
 bool Type::isSignlessInteger() const {
-  if (auto intTy = dyn_cast<IntegerType>())
+  if (auto intTy = llvm::dyn_cast<IntegerType>(*this))
     return intTy.isSignless();
   return false;
 }
 
 bool Type::isSignlessInteger(unsigned width) const {
-  if (auto intTy = dyn_cast<IntegerType>())
+  if (auto intTy = llvm::dyn_cast<IntegerType>(*this))
     return intTy.isSignless() && intTy.getWidth() == width;
   return false;
 }
 
 bool Type::isSignedInteger() const {
-  if (auto intTy = dyn_cast<IntegerType>())
+  if (auto intTy = llvm::dyn_cast<IntegerType>(*this))
     return intTy.isSigned();
   return false;
 }
 
 bool Type::isSignedInteger(unsigned width) const {
-  if (auto intTy = dyn_cast<IntegerType>())
+  if (auto intTy = llvm::dyn_cast<IntegerType>(*this))
     return intTy.isSigned() && intTy.getWidth() == width;
   return false;
 }
 
 bool Type::isUnsignedInteger() const {
-  if (auto intTy = dyn_cast<IntegerType>())
+  if (auto intTy = llvm::dyn_cast<IntegerType>(*this))
     return intTy.isUnsigned();
   return false;
 }
 
 bool Type::isUnsignedInteger(unsigned width) const {
-  if (auto intTy = dyn_cast<IntegerType>())
+  if (auto intTy = llvm::dyn_cast<IntegerType>(*this))
     return intTy.isUnsigned() && intTy.getWidth() == width;
   return false;
 }
 
 bool Type::isSignlessIntOrIndex() const {
-  return isSignlessInteger() || isa<IndexType>();
+  return isSignlessInteger() || llvm::isa<IndexType>(*this);
 }
 
 bool Type::isSignlessIntOrIndexOrFloat() const {
-  return isSignlessInteger() || isa<IndexType, FloatType>();
+  return isSignlessInteger() || llvm::isa<IndexType, FloatType>(*this);
 }
 
 bool Type::isSignlessIntOrFloat() const {
-  return isSignlessInteger() || isa<FloatType>();
+  return isSignlessInteger() || llvm::isa<FloatType>(*this);
 }
 
-bool Type::isIntOrIndex() const { return isa<IntegerType>() || isIndex(); }
+bool Type::isIntOrIndex() const {
+  return llvm::isa<IntegerType>(*this) || isIndex();
+}
 
-bool Type::isIntOrFloat() const { return isa<IntegerType, FloatType>(); }
+bool Type::isIntOrFloat() const {
+  return llvm::isa<IntegerType, FloatType>(*this);
+}
 
 bool Type::isIntOrIndexOrFloat() const { return isIntOrFloat() || isIndex(); }
 
 unsigned Type::getIntOrFloatBitWidth() const {
   assert(isIntOrFloat() && "only integers and floats have a bitwidth");
-  if (auto intType = dyn_cast<IntegerType>())
+  if (auto intType = llvm::dyn_cast<IntegerType>(*this))
     return intType.getWidth();
-  return cast<FloatType>().getWidth();
+  return llvm::cast<FloatType>(*this).getWidth();
 }

mlir/lib/IR/Unit.cpp

static void printBlock(llvm::raw_ostream &os, Block *block,
   os << "Block #" << blockId << " for ";
   bool shouldSkipRegions = flags.shouldSkipRegions();
   printRegion(os, region, flags.skipRegions());
   if (!shouldSkipRegions)
     block->print(os);
 }
 
 void mlir::IRUnit::print(llvm::raw_ostream &os, OpPrintingFlags flags) const {
-  if (auto *op = this->dyn_cast<Operation *>())
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(*this))
     return printOp(os, op, flags);
-  if (auto *region = this->dyn_cast<Region *>())
+  if (auto *region = llvm::dyn_cast_if_present<Region *>(*this))
     return printRegion(os, region, flags);
-  if (auto *block = this->dyn_cast<Block *>())
+  if (auto *block = llvm::dyn_cast_if_present<Block *>(*this))
     return printBlock(os, block, flags);
   llvm_unreachable("unknown IRUnit");
 }
 
 llvm::raw_ostream &mlir::operator<<(llvm::raw_ostream &os, const IRUnit &unit) {
   unit.print(os);
   return os;
 }

mlir/lib/IR/Value.cpp

 #include "llvm/ADT/SmallPtrSet.h"
 
 using namespace mlir;
 using namespace mlir::detail;
 
 /// If this value is the result of an Operation, return the operation that
 /// defines it.
 Operation *Value::getDefiningOp() const {
-  if (auto result = dyn_cast<OpResult>())
+  if (auto result = llvm::dyn_cast<OpResult>(*this))
     return result.getOwner();
   return nullptr;
 }
 
 Location Value::getLoc() const {
   if (auto *op = getDefiningOp())
     return op->getLoc();
 
-  return cast<BlockArgument>().getLoc();
+  return llvm::cast<BlockArgument>(*this).getLoc();
 }
 
 void Value::setLoc(Location loc) {
   if (auto *op = getDefiningOp())
     return op->setLoc(loc);
 
-  return cast<BlockArgument>().setLoc(loc);
+  return llvm::cast<BlockArgument>(*this).setLoc(loc);
 }
 
 /// Return the Region in which this Value is defined.
 Region *Value::getParentRegion() {
   if (auto *op = getDefiningOp())
     return op->getParentRegion();
-  return cast<BlockArgument>().getOwner()->getParent();
+  return llvm::cast<BlockArgument>(*this).getOwner()->getParent();
 }
 
 /// Return the Block in which this Value is defined.
 Block *Value::getParentBlock() {
   if (Operation *op = getDefiningOp())
     return op->getBlock();
-  return cast<BlockArgument>().getOwner();
+  return llvm::cast<BlockArgument>(*this).getOwner();
 }
 
 //===----------------------------------------------------------------------===//
 // Value::UseLists
 //===----------------------------------------------------------------------===//
 
 /// Replace all uses of 'this' value with the new value, updating anything in
 /// the IR that uses 'this' to use the other value instead except if the user is

mlir/lib/Interfaces/DataLayoutInterfaces.cpp

mlir::detail::getDefaultStackAlignment(DataLayoutEntryInterface entry) {
   return value.getValue().getZExtValue();
 }
 
 DataLayoutEntryList
 mlir::detail::filterEntriesForType(DataLayoutEntryListRef entries,
                                    TypeID typeID) {
   return llvm::to_vector<4>(llvm::make_filter_range(
       entries, [typeID](DataLayoutEntryInterface entry) {
-        auto type = entry.getKey().dyn_cast<Type>();
+        auto type = llvm::dyn_cast_if_present<Type>(entry.getKey());
         return type && type.getTypeID() == typeID;
       }));
 }
 
 DataLayoutEntryInterface
 mlir::detail::filterEntryForIdentifier(DataLayoutEntryListRef entries,
                                        StringAttr id) {
   const auto *it = llvm::find_if(entries, [id](DataLayoutEntryInterface entry) {

 //===----------------------------------------------------------------------===//
 // DataLayoutSpecInterface
 //===----------------------------------------------------------------------===//
 
 void DataLayoutSpecInterface::bucketEntriesByType(
     DenseMap<TypeID, DataLayoutEntryList> &types,
     DenseMap<StringAttr, DataLayoutEntryInterface> &ids) {
   for (DataLayoutEntryInterface entry : getEntries()) {
-    if (auto type = entry.getKey().dyn_cast<Type>())
+    if (auto type = llvm::dyn_cast_if_present<Type>(entry.getKey()))
       types[type.getTypeID()].push_back(entry);
     else
       ids[entry.getKey().get<StringAttr>()] = entry;
   }
 }
 
 LogicalResult mlir::detail::verifyDataLayoutSpec(DataLayoutSpecInterface spec,
                                                  Location loc) {

mlir/lib/Interfaces/InferTypeOpInterface.cpp

assert(resultIdx == static_cast<int64_t>(reifiedReturnShapes.size()) &&
          "incorrect implementation of ReifyRankedShapedTypeOpInterface");
 #endif // NDEBUG
   return status;
 }
 
 bool ShapeAdaptor::hasRank() const {
   if (val.isNull())
     return false;
-  if (auto t = val.dyn_cast<Type>())
+  if (auto t = llvm::dyn_cast_if_present<Type>(val))
     return cast<ShapedType>(t).hasRank();
   if (val.is<Attribute>())
     return true;
   return val.get<ShapedTypeComponents *>()->hasRank();
 }
 
 Type ShapeAdaptor::getElementType() const {
   if (val.isNull())
     return nullptr;
-  if (auto t = val.dyn_cast<Type>())
+  if (auto t = llvm::dyn_cast_if_present<Type>(val))
     return cast<ShapedType>(t).getElementType();
   if (val.is<Attribute>())
     return nullptr;
   return val.get<ShapedTypeComponents *>()->getElementType();
 }
 
 void ShapeAdaptor::getDims(SmallVectorImpl<int64_t> &res) const {
   assert(hasRank());
-  if (auto t = val.dyn_cast<Type>()) {
+  if (auto t = llvm::dyn_cast_if_present<Type>(val)) {
     ArrayRef<int64_t> vals = cast<ShapedType>(t).getShape();
     res.assign(vals.begin(), vals.end());
-  } else if (auto attr = val.dyn_cast<Attribute>()) {
+  } else if (auto attr = llvm::dyn_cast_if_present<Attribute>(val)) {
     auto dattr = cast<DenseIntElementsAttr>(attr);
     res.clear();
     res.reserve(dattr.size());
     for (auto it : dattr.getValues<APInt>())
       res.push_back(it.getSExtValue());
   } else {
     auto vals = val.get<ShapedTypeComponents *>()->getDims();
     res.assign(vals.begin(), vals.end());
   }
 }
 
 void ShapeAdaptor::getDims(ShapedTypeComponents &res) const {
   assert(hasRank());
   res.ranked = true;
   getDims(res.dims);
 }
 
 int64_t ShapeAdaptor::getDimSize(int index) const {
   assert(hasRank());
-  if (auto t = val.dyn_cast<Type>())
+  if (auto t = llvm::dyn_cast_if_present<Type>(val))
     return cast<ShapedType>(t).getDimSize(index);
-  if (auto attr = val.dyn_cast<Attribute>())
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(val))
     return cast<DenseIntElementsAttr>(attr)
         .getValues<APInt>()[index]
         .getSExtValue();
   auto *stc = val.get<ShapedTypeComponents *>();
   return stc->getDims()[index];
 }
 
 int64_t ShapeAdaptor::getRank() const {
   assert(hasRank());
-  if (auto t = val.dyn_cast<Type>())
+  if (auto t = llvm::dyn_cast_if_present<Type>(val))
     return cast<ShapedType>(t).getRank();
-  if (auto attr = val.dyn_cast<Attribute>())
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(val))
     return cast<DenseIntElementsAttr>(attr).size();
   return val.get<ShapedTypeComponents *>()->getDims().size();
 }
 
 bool ShapeAdaptor::hasStaticShape() const {
   if (!hasRank())
     return false;
 
-  if (auto t = val.dyn_cast<Type>())
+  if (auto t = llvm::dyn_cast_if_present<Type>(val))
     return cast<ShapedType>(t).hasStaticShape();
-  if (auto attr = val.dyn_cast<Attribute>()) {
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(val)) {
     auto dattr = cast<DenseIntElementsAttr>(attr);
     for (auto index : dattr.getValues<APInt>())
       if (ShapedType::isDynamic(index.getSExtValue()))
         return false;
     return true;
   }
   auto *stc = val.get<ShapedTypeComponents *>();
   return llvm::none_of(stc->getDims(), ShapedType::isDynamic);
 }
 
 int64_t ShapeAdaptor::getNumElements() const {
   assert(hasStaticShape() && "cannot get element count of dynamic shaped type");
 
-  if (auto t = val.dyn_cast<Type>())
+  if (auto t = llvm::dyn_cast_if_present<Type>(val))
     return cast<ShapedType>(t).getNumElements();
 
-  if (auto attr = val.dyn_cast<Attribute>()) {
+  if (auto attr = llvm::dyn_cast_if_present<Attribute>(val)) {
     auto dattr = cast<DenseIntElementsAttr>(attr);
     int64_t num = 1;
     for (auto index : dattr.getValues<APInt>()) {
       num *= index.getZExtValue();
       assert(num >= 0 && "integer overflow in element count computation");
     }
     return num;
   }

mlir/lib/Interfaces/ValueBoundsOpInterface.cpp

 
 namespace mlir {
 #include "mlir/Interfaces/ValueBoundsOpInterface.cpp.inc"
 } // namespace mlir
 
 /// If ofr is a constant integer or an IntegerAttr, return the integer.
 static std::optional<int64_t> getConstantIntValue(OpFoldResult ofr) {
   // Case 1: Check for Constant integer.
-  if (auto val = ofr.dyn_cast<Value>()) {
+  if (auto val = llvm::dyn_cast_if_present<Value>(ofr)) {
     APSInt intVal;
     if (matchPattern(val, m_ConstantInt(&intVal)))
       return intVal.getSExtValue();
     return std::nullopt;
   }
   // Case 2: Check for IntegerAttr.
-  Attribute attr = ofr.dyn_cast<Attribute>();
+  Attribute attr = llvm::dyn_cast_if_present<Attribute>(ofr);
   if (auto intAttr = dyn_cast_or_null<IntegerAttr>(attr))
     return intAttr.getValue().getSExtValue();
   return std::nullopt;
 }
 
 ValueBoundsConstraintSet::ValueBoundsConstraintSet(MLIRContext *ctx)
     : builder(ctx) {}
 
if (!valueDimToPosition.contains(valueDim))
     (void)insert(value, dim);
   int64_t pos = getPos(value, dim);
   return pos < cstr.getNumDimVars()
              ? builder.getAffineDimExpr(pos)
              : builder.getAffineSymbolExpr(pos - cstr.getNumDimVars());
 }
 
 AffineExpr ValueBoundsConstraintSet::getExpr(OpFoldResult ofr) {
-  if (Value value = ofr.dyn_cast<Value>())
+  if (Value value = llvm::dyn_cast_if_present<Value>(ofr))
     return getExpr(value, /*dim=*/std::nullopt);
   auto constInt = getConstantIntValue(ofr);
   assert(constInt.has_value() && "expected Integer constant");
   return builder.getAffineConstantExpr(*constInt);
 }
 
 AffineExpr ValueBoundsConstraintSet::getExpr(int64_t constant) {
   return builder.getAffineConstantExpr(constant);

mlir/lib/Pass/PassDetail.h

struct PassExecutionAction : public tracing::ActionImpl<PassExecutionAction> {
   using Base = tracing::ActionImpl<PassExecutionAction>;
   PassExecutionAction(ArrayRef<IRUnit> irUnits, const Pass &pass)
       : Base(irUnits), pass(pass) {}
   static constexpr StringLiteral tag = "pass-execution";
   void print(raw_ostream &os) const override;
   const Pass &getPass() const { return pass; }
   Operation *getOp() const {
     ArrayRef<IRUnit> irUnits = getContextIRUnits();
-    return irUnits.empty() ? nullptr : irUnits[0].dyn_cast<Operation *>();
+    return irUnits.empty() ? nullptr
+                           : llvm::dyn_cast_if_present<Operation *>(irUnits[0]);
   }
 
 public:
   const Pass &pass;
   Operation *op;
 };
 
 namespace detail {

mlir/lib/TableGen/Operator.cpp

void Operator::populateTypeInferenceInfo(
   if (getNumVariableLengthResults() > 0)
     return;
 
   // Skip cases currently being custom generated.
   // TODO: Remove special cases.
   if (getTrait("::mlir::OpTrait::SameOperandsAndResultType")) {
     // Check for a non-variable length operand to use as the type anchor.
     auto *operandI = llvm::find_if(arguments, [](const Argument &arg) {
-      NamedTypeConstraint *operand = arg.dyn_cast<NamedTypeConstraint *>();
+      NamedTypeConstraint *operand = llvm::dyn_cast_if_present<NamedTypeConstraint *>(arg);
       return operand && !operand->isVariableLength();
     });
     if (operandI == arguments.end())
       return;
 
     // All result types are inferred from the operand type.
     int operandIdx = operandI - arguments.begin();
     for (int i = 0; i < getNumResults(); ++i)

 StringRef Operator::getAssemblyFormat() const {
   return TypeSwitch<llvm::Init *, StringRef>(def.getValueInit("assemblyFormat"))
       .Case<llvm::StringInit>([&](auto *init) { return init->getValue(); });
 }
 
 void Operator::print(llvm::raw_ostream &os) const {
   os << "op '" << getOperationName() << "'\n";
   for (Argument arg : arguments) {
-    if (auto *attr = arg.dyn_cast<NamedAttribute *>())
+    if (auto *attr = llvm::dyn_cast_if_present<NamedAttribute *>(arg))
       os << "[attribute] " << attr->name << '\n';
     else
       os << "[operand] " << arg.get<NamedTypeConstraint *>()->name << '\n';
   }
 }
 
 auto Operator::VariableDecoratorIterator::unwrap(llvm::Init *init)
     -> VariableDecorator {

mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp

 /// otherwise returns nullptr.
 static llvm::MDNode *
 convertBranchWeights(std::optional<ElementsAttr> weights,
                      LLVM::ModuleTranslation &moduleTranslation) {
   if (!weights)
     return nullptr;
   SmallVector<uint32_t> weightValues;
   weightValues.reserve(weights->size());
-  for (APInt weight : weights->cast<DenseIntElementsAttr>())
+  for (APInt weight : llvm::cast<DenseIntElementsAttr>(*weights))
     weightValues.push_back(weight.getLimitedValue());
   return llvm::MDBuilder(moduleTranslation.getLLVMContext())
       .createBranchWeights(weightValues);
 }
 
 static LogicalResult
 convertOperationImpl(Operation &opInst, llvm::IRBuilderBase &builder,
                      LLVM::ModuleTranslation &moduleTranslation) {
if (auto switchOp = dyn_cast<LLVM::SwitchOp>(opInst)) {
     llvm::SwitchInst *switchInst = builder.CreateSwitch(
         moduleTranslation.lookupValue(switchOp.getValue()),
         moduleTranslation.lookupBlock(switchOp.getDefaultDestination()),
         switchOp.getCaseDestinations().size(), branchWeights);
 
     auto *ty = llvm::cast<llvm::IntegerType>(
         moduleTranslation.convertType(switchOp.getValue().getType()));
     for (auto i :
-         llvm::zip(switchOp.getCaseValues()->cast<DenseIntElementsAttr>(),
+         llvm::zip(llvm::cast<DenseIntElementsAttr>(*switchOp.getCaseValues()),
                    switchOp.getCaseDestinations()))
       switchInst->addCase(
           llvm::ConstantInt::get(ty, std::get<0>(i).getLimitedValue()),
           moduleTranslation.lookupBlock(std::get<1>(i)));
 
     moduleTranslation.mapBranch(&opInst, switchInst);
     return success();
   }

mlir/lib/Target/LLVMIR/ModuleImport.cpp

if (Attribute scalarAttr = getScalarConstantAsAttr(builder, constant))
     return scalarAttr;
 
   // Convert function references.
   if (auto *func = dyn_cast<llvm::Function>(constant))
     return SymbolRefAttr::get(builder.getContext(), func->getName());
 
   // Returns the static shape of the provided type if possible.
   auto getConstantShape = [&](llvm::Type *type) {
-    return getBuiltinTypeForAttr(convertType(type))
-        .dyn_cast_or_null<ShapedType>();
+    return llvm::dyn_cast_if_present<ShapedType>(getBuiltinTypeForAttr(convertType(type))
+        );
   };
 
   // Convert one-dimensional constant arrays or vectors that store 1/2/4/8-byte
   // integer or half/bfloat/float/double values.
   if (auto *constArray = dyn_cast<llvm::ConstantDataSequential>(constant)) {
     if (constArray->isString())
       return builder.getStringAttr(constArray->getAsString());
     auto shape = getConstantShape(constArray->getType());
while (!workList.empty()) {
       // constant expression.
       return {};
     }
     return DenseElementsAttr::get(shape, elementAttrs);
   }
 
   // Convert zero aggregates.
   if (auto *constZero = dyn_cast<llvm::ConstantAggregateZero>(constant)) {
-    auto shape = getBuiltinTypeForAttr(convertType(constZero->getType()))
-                     .dyn_cast_or_null<ShapedType>();
+    auto shape = llvm::dyn_cast_if_present<ShapedType>(getBuiltinTypeForAttr(convertType(constZero->getType()))
+                     );
     if (!shape)
       return {};
     // Convert zero aggregates with a static shape to splat elements attributes.
     Attribute splatAttr = builder.getZeroAttr(shape.getElementType());
     assert(splatAttr && "expected non-null zero attribute for scalar types");
     return SplatElementsAttr::get(shape, splatAttr);
   }
   return {};

mlir/lib/Target/LLVMIR/ModuleTranslation.cpp

translateDataLayout(DataLayoutSpecInterface attribute,
                     std::optional<Location> loc = std::nullopt) {
   if (!loc)
     loc = UnknownLoc::get(attribute.getContext());
 
   // Translate the endianness attribute.
   std::string llvmDataLayout;
   llvm::raw_string_ostream layoutStream(llvmDataLayout);
   for (DataLayoutEntryInterface entry : attribute.getEntries()) {
-    auto key = entry.getKey().dyn_cast<StringAttr>();
+    auto key = llvm::dyn_cast_if_present<StringAttr>(entry.getKey());
     if (!key)
       continue;
     if (key.getValue() == DLTIDialect::kDataLayoutEndiannessKey) {
       auto value = cast<StringAttr>(entry.getValue());
       bool isLittleEndian =
           value.getValue() == DLTIDialect::kDataLayoutEndiannessLittle;
       layoutStream << "-" << (isLittleEndian ? "e" : "E");
       layoutStream.flush();
for (DataLayoutEntryInterface entry : attribute.getEntries()) {
     emitError(*loc) << "unsupported data layout key " << key;
     return failure();
   }
 
   // Go through the list of entries to check which types are explicitly
   // specified in entries. Where possible, data layout queries are used instead
   // of directly inspecting the entries.
   for (DataLayoutEntryInterface entry : attribute.getEntries()) {
-    auto type = entry.getKey().dyn_cast<Type>();
+    auto type = llvm::dyn_cast_if_present<Type>(entry.getKey());
     if (!type)
       continue;
     // Data layout for the index type is irrelevant at this point.
     if (isa<IndexType>(type))
       continue;
     layoutStream << "-";
     LogicalResult result =
         llvm::TypeSwitch<Type, LogicalResult>(type)

mlir/lib/Target/SPIRV/Deserialization/Deserializer.cpp

case spirv::Decoration::LinkageAttributes: {
     // being variable length, words.size() = 3 + strlen(name)/4 + 1 or
     // 3 + ceildiv(strlen(name), 4).
     unsigned wordIndex = 2;
     auto linkageName = spirv::decodeStringLiteral(words, wordIndex).str();
     auto linkageTypeAttr = opBuilder.getAttr<::mlir::spirv::LinkageTypeAttr>(
         static_cast<::mlir::spirv::LinkageType>(words[wordIndex++]));
     auto linkageAttr = opBuilder.getAttr<::mlir::spirv::LinkageAttributesAttr>(
         linkageName, linkageTypeAttr);
-    decorations[words[0]].set(symbol, linkageAttr.dyn_cast<Attribute>());
+    decorations[words[0]].set(symbol, llvm::dyn_cast<Attribute>(linkageAttr));
     break;
   }
   case spirv::Decoration::Aliased:
   case spirv::Decoration::Block:
   case spirv::Decoration::BufferBlock:
   case spirv::Decoration::Flat:
   case spirv::Decoration::NonReadable:
   case spirv::Decoration::NonWritable:

mlir/lib/Target/SPIRV/Serialization/SerializeOps.cpp

Serializer::processOp<spirv::ExecutionModeOp>(spirv::ExecutionModeOp op) {
   // Add the ExecutionMode.
   operands.push_back(static_cast<uint32_t>(op.getExecutionMode()));
 
   // Serialize values if any.
   auto values = op.getValues();
   if (values) {
     for (auto &intVal : values.getValue()) {
       operands.push_back(static_cast<uint32_t>(
-          intVal.cast<IntegerAttr>().getValue().getZExtValue()));
+          llvm::cast<IntegerAttr>(intVal).getValue().getZExtValue()));
     }
   }
   encodeInstructionInto(executionModes, spirv::Opcode::OpExecutionMode,
                         operands);
   return success();
 }
 
 template <>

mlir/lib/Target/SPIRV/Serialization/Serializer.cpp

return emitError(
                     "decoration name, unhandled attribute with name : ")
            << attrName;
   }
   SmallVector<uint32_t, 1> args;
   switch (*decoration) {
   case spirv::Decoration::LinkageAttributes: {
     // Get the value of the Linkage Attributes
     // e.g., LinkageAttributes=["linkageName", linkageType].
-    auto linkageAttr = attr.getValue().dyn_cast<spirv::LinkageAttributesAttr>();
+    auto linkageAttr = llvm::dyn_cast<spirv::LinkageAttributesAttr>(attr.getValue());
     auto linkageName = linkageAttr.getLinkageName();
     auto linkageType = linkageAttr.getLinkageType().getValue();
     // Encode the Linkage Name (string literal to uint32_t).
     spirv::encodeStringLiteralInto(args, linkageName);
     // Encode LinkageType & Add the Linkagetype to the args.
     args.push_back(static_cast<uint32_t>(linkageType));
     break;
   }

mlir/lib/Tools/mlir-pdll-lsp-server/PDLLServer.cpp

 struct PDLIndexSymbol {
   explicit PDLIndexSymbol(const ast::Decl *definition)
       : definition(definition) {}
   explicit PDLIndexSymbol(const ods::Operation *definition)
       : definition(definition) {}
 
   /// Return the location of the definition of this symbol.
   SMRange getDefLoc() const {
-    if (const ast::Decl *decl = definition.dyn_cast<const ast::Decl *>()) {
+    if (const ast::Decl *decl = llvm::dyn_cast_if_present<const ast::Decl *>(definition)) {
       const ast::Name *declName = decl->getName();
       return declName ? declName->getLoc() : decl->getLoc();
     }
     return definition.get<const ods::Operation *>()->getLoc();
   }
 
   /// The main definition of the symbol.
   PointerUnion<const ast::Decl *, const ods::Operation *> definition;
PDLDocument::findHover(const lsp::URIForFile &uri,
 
   // Find the symbol at the given location.
   SMRange hoverRange;
   const PDLIndexSymbol *symbol = index.lookup(posLoc, &hoverRange);
   if (!symbol)
     return std::nullopt;
 
   // Add hover for operation names.
-  if (const auto *op = symbol->definition.dyn_cast<const ods::Operation *>())
+  if (const auto *op = llvm::dyn_cast_if_present<const ods::Operation *>(symbol->definition))
     return buildHoverForOpName(op, hoverRange);
   const auto *decl = symbol->definition.get<const ast::Decl *>();
   return findHover(decl, hoverRange);
 }
 
 std::optional<lsp::Hover> PDLDocument::findHover(const ast::Decl *decl,
                                                  const SMRange &hoverRange) {
   // Add hover for variables.

mlir/lib/Transforms/Inliner.cpp

 }
 
 //===----------------------------------------------------------------------===//
 // Inliner
 //===----------------------------------------------------------------------===//
 
 #ifndef NDEBUG
 static std::string getNodeName(CallOpInterface op) {
-  if (auto sym = op.getCallableForCallee().dyn_cast<SymbolRefAttr>())
+  if (auto sym = llvm::dyn_cast_if_present<SymbolRefAttr>(op.getCallableForCallee()))
     return debugString(op);
   return "_unnamed_callee_";
 }
 #endif
 
 /// Return true if the specified `inlineHistoryID`  indicates an inline history
 /// that already includes `node`.
 static bool inlineHistoryIncludes(

mlir/lib/Transforms/Utils/FoldUtils.cpp

OperationFolder::processFoldResults(Operation *op,
   auto &uniquedConstants = foldScopes[insertRegion];
 
   // Create the result constants and replace the results.
   auto *dialect = op->getDialect();
   for (unsigned i = 0, e = op->getNumResults(); i != e; ++i) {
     assert(!foldResults[i].isNull() && "expected valid OpFoldResult");
 
     // Check if the result was an SSA value.
-    if (auto repl = foldResults[i].dyn_cast<Value>()) {
+    if (auto repl = llvm::dyn_cast_if_present<Value>(foldResults[i])) {
       if (repl.getType() != op->getResult(i).getType()) {
         results.clear();
         return failure();
       }
       results.emplace_back(repl);
       continue;
     }
 

mlir/lib/Transforms/Utils/InliningUtils.cpp

inlineRegionImpl(InlinerInterface &interface, Region *src, Block *inlineBlock,
 
   // Get the range of newly inserted blocks.
   auto newBlocks = llvm::make_range(std::next(inlineBlock->getIterator()),
                                     postInsertBlock->getIterator());
   Block *firstNewBlock = &*newBlocks.begin();
 
   // Remap the locations of the inlined operations if a valid source location
   // was provided.
-  if (inlineLoc && !inlineLoc->isa<UnknownLoc>())
+  if (inlineLoc && !llvm::isa<UnknownLoc>(*inlineLoc))
     remapInlinedLocations(newBlocks, *inlineLoc);
 
   // If the blocks were moved in-place, make sure to remap any necessary
   // operands.
   if (!shouldCloneInlinedRegion)
     remapInlinedOperands(newBlocks, mapper);
 
   // Process the newly inlined blocks.

mlir/test/lib/Analysis/TestDataFlowFramework.cpp

for (Operation &op : block) {
         return op.emitError("unexpected op with regions");
       visitOperation(&op);
     }
   }
   return success();
 }
 
 LogicalResult FooAnalysis::visit(ProgramPoint point) {
-  if (auto *op = point.dyn_cast<Operation *>()) {
+  if (auto *op = llvm::dyn_cast_if_present<Operation *>(point)) {
     visitOperation(op);
     return success();
   }
-  if (auto *block = point.dyn_cast<Block *>()) {
+  if (auto *block = llvm::dyn_cast_if_present<Block *>(point)) {
     visitBlock(block);
     return success();
   }
   return emitError(point.getLoc(), "unknown point kind");
 }
 
 void FooAnalysis::visitBlock(Block *block) {
   if (block->isEntryBlock()) {

mlir/test/lib/Dialect/Affine/TestReifyValueBounds.cpp

if (op->getName().getStringRef() == "test.reify_bound" ||
         }
       }
       if (failed(reified)) {
         op->emitOpError("could not reify bound");
         return WalkResult::interrupt();
       }
 
       // Replace the op with the reified bound.
-      if (auto val = reified->dyn_cast<Value>()) {
+      if (auto val = llvm::dyn_cast_if_present<Value>(*reified)) {
         rewriter.replaceOp(op, val);
         return WalkResult::skip();
       }
       Value constOp = rewriter.create<arith::ConstantIndexOp>(
           op->getLoc(), cast<IntegerAttr>(reified->get<Attribute>()).getInt());
       rewriter.replaceOp(op, constOp);
       return WalkResult::skip();
     }

mlir/tools/mlir-tblgen/OpDefinitionsGen.cpp

const char *propFromAttrFmt = R"decl(;
                    "Properties.";
         return failure();
       }
       if (failed(setFromAttr(prop.{1}, attr, diag))) return ::mlir::failure();
     }
 )decl";
   for (const auto &attrOrProp : attrOrProperties) {
     if (const auto *namedProperty =
-            attrOrProp.dyn_cast<const NamedProperty *>()) {
+            llvm::dyn_cast_if_present<const NamedProperty *>(attrOrProp)) {
       StringRef name = namedProperty->name;
       auto &prop = namedProperty->prop;
       FmtContext fctx;
       setPropMethod << formatv(propFromAttrFmt,
                                tgfmt(prop.getConvertFromAttributeCall(),
                                      &fctx.addSubst("_attr", propertyAttr)
                                           .addSubst("_storage", propertyStorage)
                                           .addSubst("_diag", propertyDiag)),
                                name);
     } else {
-      const auto *namedAttr = attrOrProp.dyn_cast<const AttributeMetadata *>();
+      const auto *namedAttr = llvm::dyn_cast_if_present<const AttributeMetadata *>(attrOrProp);
       StringRef name = namedAttr->attrName;
       setPropMethod << formatv(R"decl(
   {{
     auto &propStorage = prop.{0};
     auto attr = dict.get("{0}");
     if (attr || /*isRequired=*/{1}) {{
       if (!attr) {{
         if (diag)
const char *propToAttrFmt = R"decl(
     {
       const auto &propStorage = prop.{0};
       attrs.push_back(odsBuilder.getNamedAttr("{0}",
                                               {1}));
     }
 )decl";
   for (const auto &attrOrProp : attrOrProperties) {
     if (const auto *namedProperty =
-            attrOrProp.dyn_cast<const NamedProperty *>()) {
+            llvm::dyn_cast_if_present<const NamedProperty *>(attrOrProp)) {
       StringRef name = namedProperty->name;
       auto &prop = namedProperty->prop;
       FmtContext fctx;
       getPropMethod << formatv(
           propToAttrFmt, name,
           tgfmt(prop.getConvertToAttributeCall(),
                 &fctx.addSubst("_ctxt", "ctx")
                      .addSubst("_storage", propertyStorage)));
       continue;
     }
-    const auto *namedAttr = attrOrProp.dyn_cast<const AttributeMetadata *>();
+    const auto *namedAttr = llvm::dyn_cast_if_present<const AttributeMetadata *>(attrOrProp);
     StringRef name = namedAttr->attrName;
     getPropMethod << formatv(R"decl(
     {{
       const auto &propStorage = prop.{0};
       if (propStorage)
         attrs.push_back(odsBuilder.getNamedAttr("{0}",
                                        propStorage));
     }
)decl";
 
   const char *propHashFmt = R"decl(
   auto hash_{0} = [] (const auto &propStorage) -> llvm::hash_code {
     return {1};
   };
 )decl";
   for (const auto &attrOrProp : attrOrProperties) {
     if (const auto *namedProperty =
-            attrOrProp.dyn_cast<const NamedProperty *>()) {
+            llvm::dyn_cast_if_present<const NamedProperty *>(attrOrProp)) {
       StringRef name = namedProperty->name;
       auto &prop = namedProperty->prop;
       FmtContext fctx;
       hashMethod << formatv(propHashFmt, name,
                             tgfmt(prop.getHashPropertyCall(),
                                   &fctx.addSubst("_storage", propertyStorage)));
     }
   }
   hashMethod << "  return llvm::hash_combine(";
   llvm::interleaveComma(
       attrOrProperties, hashMethod, [&](const ConstArgument &attrOrProp) {
         if (const auto *namedProperty =
-                attrOrProp.dyn_cast<const NamedProperty *>()) {
+                llvm::dyn_cast_if_present<const NamedProperty *>(attrOrProp)) {
           hashMethod << "\n    hash_" << namedProperty->name << "(prop."
                      << namedProperty->name << ")";
           return;
         }
         const auto *namedAttr =
-            attrOrProp.dyn_cast<const AttributeMetadata *>();
+            llvm::dyn_cast_if_present<const AttributeMetadata *>(attrOrProp);
         StringRef name = namedAttr->attrName;
         hashMethod << "\n    llvm::hash_value(prop." << name
                    << ".getAsOpaquePointer())";
       });
   hashMethod << ");\n";
 
   const char *getInherentAttrMethodFmt = R"decl(
     if (name == "{0}")
       return prop.{0};
 )decl";
   const char *setInherentAttrMethodFmt = R"decl(
     if (name == "{0}") {{
        prop.{0} = dyn_cast_or_null<std::remove_reference_t<decltype(prop.{0})>>(value);
        return;
     }
 )decl";
   const char *populateInherentAttrsMethodFmt = R"decl(
     if (prop.{0}) attrs.append("{0}", prop.{0});
 )decl";
   for (const auto &attrOrProp : attrOrProperties) {
     if (const auto *namedAttr =
-            attrOrProp.dyn_cast<const AttributeMetadata *>()) {
+            llvm::dyn_cast_if_present<const AttributeMetadata *>(attrOrProp)) {
       StringRef name = namedAttr->attrName;
       getInherentAttrMethod << formatv(getInherentAttrMethodFmt, name);
       setInherentAttrMethod << formatv(setInherentAttrMethodFmt, name);
       populateInherentAttrsMethod
           << formatv(populateInherentAttrsMethodFmt, name);
       continue;
     }
   }
   getInherentAttrMethod << "  return std::nullopt;\n";
 
   // Emit the verifiers method for backward compatibility with the generic
   // syntax. This method verifies the constraint on the properties attributes
   // before they are set, since dyn_cast<> will silently omit failures.
   for (const auto &attrOrProp : attrOrProperties) {
-    const auto *namedAttr = attrOrProp.dyn_cast<const AttributeMetadata *>();
+    const auto *namedAttr = llvm::dyn_cast_if_present<const AttributeMetadata *>(attrOrProp);
     if (!namedAttr || !namedAttr->constraint)
       continue;
     Attribute attr = *namedAttr->constraint;
     std::optional<StringRef> constraintFn =
         staticVerifierEmitter.getAttrConstraintFn(attr);
     if (!constraintFn)
       continue;
     if (canEmitAttrVerifier(attr,
void OpEmitter::buildParamList(SmallVectorImpl<MethodParameter> &paramList,
   int defaultValuedAttrStartIndex = op.getNumArgs();
   // Successors and variadic regions go at the end of the parameter list, so no
   // default arguments are possible.
   bool hasTrailingParams = op.getNumSuccessors() || op.getNumVariadicRegions();
   if (attrParamKind == AttrParamKind::UnwrappedValue && !hasTrailingParams) {
     // Calculate the start index from which we can attach default values in the
     // builder declaration.
     for (int i = op.getNumArgs() - 1; i >= 0; --i) {
-      auto *namedAttr = op.getArg(i).dyn_cast<tblgen::NamedAttribute *>();
+      auto *namedAttr = llvm::dyn_cast_if_present<tblgen::NamedAttribute *>(op.getArg(i));
       if (!namedAttr || !namedAttr->attr.hasDefaultValue())
         break;
 
       if (!canUseUnwrappedRawValue(namedAttr->attr))
         break;
 
       // Creating an APInt requires us to provide bitwidth, value, and
       // signedness, which is complicated compared to others. Similarly
for (const NamedTypeConstraint &operand : op.getOperands()) {
     if (operand.isVariadicOfVariadic()) {
       inferredAttributes.insert(
           operand.constraint.getVariadicOfVariadicSegmentSizeAttr());
     }
   }
 
   for (int i = 0, e = op.getNumArgs(), numOperands = 0; i < e; ++i) {
     Argument arg = op.getArg(i);
-    if (const auto *operand = arg.dyn_cast<NamedTypeConstraint *>()) {
+    if (const auto *operand = llvm::dyn_cast_if_present<NamedTypeConstraint *>(arg)) {
       StringRef type;
       if (operand->isVariadicOfVariadic())
         type = "::llvm::ArrayRef<::mlir::ValueRange>";
       else if (operand->isVariadic())
         type = "::mlir::ValueRange";
       else
         type = "::mlir::Value";
 
       paramList.emplace_back(type, getArgumentName(op, numOperands++),
                              operand->isOptional());
       continue;
     }
-    if (const auto *operand = arg.dyn_cast<NamedProperty *>()) {
+    if (const auto *operand = llvm::dyn_cast_if_present<NamedProperty *>(arg)) {
       // TODO
       continue;
     }
     const NamedAttribute &namedAttr = *arg.get<NamedAttribute *>();
     const Attribute &attr = namedAttr.attr;
 
     // Inferred attributes don't need to be added to the param list.
     if (inferredAttributes.contains(namedAttr.name))
if (useProperties) {
     std::string declarations = "  struct Properties {\n";
     llvm::raw_string_ostream os(declarations);
     std::string comparator =
         "    bool operator==(const Properties &rhs) const {\n"
         "      return \n";
     llvm::raw_string_ostream comparatorOs(comparator);
     for (const auto &attrOrProp : attrOrProperties) {
       if (const auto *namedProperty =
-              attrOrProp.dyn_cast<const NamedProperty *>()) {
+              llvm::dyn_cast_if_present<const NamedProperty *>(attrOrProp)) {
         StringRef name = namedProperty->name;
         if (name.empty())
           report_fatal_error("missing name for property");
         std::string camelName =
             convertToCamelFromSnakeCase(name, /*capitalizeFirst=*/true);
         auto &prop = namedProperty->prop;
         // Generate the data member using the storage type.
         os << "    using " << name << "Ty = " << prop.getStorageType() << ";\n"
)decl";
         os << formatv(accessorFmt, prop.getInterfaceType(), camelName, name,
                       tgfmt(prop.getConvertFromStorageCall(),
                             &fctx.addSubst("_storage", propertyStorage)),
                       tgfmt(prop.getAssignToStorageCall(),
                             &fctx.addSubst("_value", propertyValue)
                                  .addSubst("_storage", propertyStorage)));
         continue;
       }
-      const auto *namedAttr = attrOrProp.dyn_cast<const AttributeMetadata *>();
+      const auto *namedAttr = llvm::dyn_cast_if_present<const AttributeMetadata *>(attrOrProp);
       const Attribute *attr = nullptr;
       if (namedAttr->constraint)
         attr = &*namedAttr->constraint;
       StringRef name = namedAttr->attrName;
       if (name.empty())
         report_fatal_error("missing name for property attr");
       std::string camelName =
           convertToCamelFromSnakeCase(name, /*capitalizeFirst=*/true);

mlir/tools/mlir-tblgen/OpFormatGen.cpp

public:
     TypeResolution() = default;
 
     /// Get the index into the buildable types for this type, or std::nullopt.
     std::optional<int> getBuilderIdx() const { return builderIdx; }
     void setBuilderIdx(int idx) { builderIdx = idx; }
 
     /// Get the variable this type is resolved to, or nullptr.
     const NamedTypeConstraint *getVariable() const {
-      return resolver.dyn_cast<const NamedTypeConstraint *>();
+      return llvm::dyn_cast_if_present<const NamedTypeConstraint *>(resolver);
     }
     /// Get the attribute this type is resolved to, or nullptr.
     const NamedAttribute *getAttribute() const {
-      return resolver.dyn_cast<const NamedAttribute *>();
+      return llvm::dyn_cast_if_present<const NamedAttribute *>(resolver);
     }
     /// Get the transformer for the type of the variable, or std::nullopt.
     std::optional<StringRef> getVarTransformer() const {
       return variableTransformer;
     }
     void setResolver(ConstArgument arg, std::optional<StringRef> transformer) {
       resolver = arg;
       variableTransformer = transformer;

mlir/tools/mlir-tblgen/OpPythonBindingGen.cpp

 /// field.
 static void
 populateBuilderLinesAttr(const Operator &op,
                          llvm::ArrayRef<std::string> argNames,
                          llvm::SmallVectorImpl<std::string> &builderLines) {
   builderLines.push_back("_ods_context = _ods_get_default_loc_context(loc)");
   for (int i = 0, e = op.getNumArgs(); i < e; ++i) {
     Argument arg = op.getArg(i);
-    auto *attribute = arg.dyn_cast<NamedAttribute *>();
+    auto *attribute = llvm::dyn_cast_if_present<NamedAttribute *>(arg);
     if (!attribute)
       continue;
 
     // Unit attributes are handled specially.
     if (attribute->attr.getStorageType().trim().equals("::mlir::UnitAttr")) {
       builderLines.push_back(llvm::formatv(initUnitAttributeTemplate,
                                            attribute->name, argNames[i]));
       continue;
auto isKeywordArgFn = [&](size_t builderArgIndex) -> bool {
     if ((builderArgIndex < numResultArgs) ||
         (builderArgIndex >= (numResultArgs + numOperandAttrArgs)))
       return false;
     // Keyword arguments:
     // - optional named attributes (including unit attributes)
     // - default-valued named attributes
     // - optional operands
     Argument a = op.getArg(builderArgIndex - numResultArgs);
-    if (auto *nattr = a.dyn_cast<NamedAttribute *>())
+    if (auto *nattr = llvm::dyn_cast_if_present<NamedAttribute *>(a))
       return (nattr->attr.isOptional() || nattr->attr.hasDefaultValue());
-    if (auto *ntype = a.dyn_cast<NamedTypeConstraint *>())
+    if (auto *ntype = llvm::dyn_cast_if_present<NamedTypeConstraint *>(a))
       return ntype->isOptional();
     return false;
   };
 
   // StringRefs in functionArgs refer to strings allocated by builderArgs.
   llvm::SmallVector<llvm::StringRef> functionArgs;
 
   // Add positional arguments.

mlir/tools/mlir-tblgen/RewriterGen.cpp

for (int i = 0, opArgIdx = 0, e = tree.getNumArgs(), nextOperand = 0; i != e;
     // Handle nested DAG construct first
     if (DagNode argTree = tree.getArgAsNestedDag(i)) {
       if (argTree.isEither()) {
         emitEitherOperandMatch(tree, argTree, castedName, opArgIdx, nextOperand,
                                depth);
         ++opArgIdx;
         continue;
       }
-      if (auto *operand = opArg.dyn_cast<NamedTypeConstraint *>()) {
+      if (auto *operand = llvm::dyn_cast_if_present<NamedTypeConstraint *>(opArg)) {
         if (operand->isVariableLength()) {
           auto error = formatv("use nested DAG construct to match op {0}'s "
                                "variadic operand #{1} unsupported now",
                                op.getOperationName(), opArgIdx);
           PrintFatalError(loc, error);
         }
       }
 
void PatternEmitter::createSeparateLocalVarsForOpArgs(
   // Now prepare operands used for building this op:
   // * If the operand is non-variadic, we create a `Value` local variable.
   // * If the operand is variadic, we create a `SmallVector<Value>` local
   //   variable.
 
   int valueIndex = 0; // An index for uniquing local variable names.
   for (int argIndex = 0, e = resultOp.getNumArgs(); argIndex < e; ++argIndex) {
     const auto *operand =
-        resultOp.getArg(argIndex).dyn_cast<NamedTypeConstraint *>();
+        llvm::dyn_cast_if_present<NamedTypeConstraint *>(resultOp.getArg(argIndex));
     // We do not need special handling for attributes.
     if (!operand)
       continue;
 
     raw_indented_ostream::DelimitedScope scope(os);
     std::string varName;
     if (operand->isVariadic()) {
       varName = std::string(formatv("tblgen_values_{0}", valueIndex++));
void PatternEmitter::supplyValuesForOpArgs(
   Operator &resultOp = node.getDialectOp(opMap);
   for (int argIndex = 0, numOpArgs = resultOp.getNumArgs();
        argIndex != numOpArgs; ++argIndex) {
     // Start each argument on its own line.
     os << ",\n    ";
 
     Argument opArg = resultOp.getArg(argIndex);
     // Handle the case of operand first.
-    if (auto *operand = opArg.dyn_cast<NamedTypeConstraint *>()) {
+    if (auto *operand = llvm::dyn_cast_if_present<NamedTypeConstraint *>(opArg)) {
       if (!operand->name.empty())
         os << "/*" << operand->name << "=*/";
       os << childNodeNames.lookup(argIndex);
       continue;
     }
 
     // The argument in the op definition.
     auto opArgName = resultOp.getArgName(argIndex);

mlir/tools/mlir-tblgen/SPIRVUtilsGen.cpp

 /// operand(attribute) is added to the `operands` list or `attributes` list.
 static void emitOperandDeserialization(const Operator &op, ArrayRef<SMLoc> loc,
                                        StringRef tabs, StringRef words,
                                        StringRef wordIndex, StringRef operands,
                                        StringRef attributes, raw_ostream &os) {
   // Process operands/attributes
   for (unsigned i = 0, e = op.getNumArgs(); i < e; ++i) {
     auto argument = op.getArg(i);
-    if (auto *valueArg = argument.dyn_cast<NamedTypeConstraint *>()) {
+    if (auto *valueArg = llvm::dyn_cast_if_present<NamedTypeConstraint *>(argument)) {
       if (valueArg->isVariableLength()) {
         if (i != e - 1) {
           PrintFatalError(loc, "SPIR-V ops can have Variadic<..> or "
                                "std::optional<...> arguments only if "
                                "it's the last argument");
         }
         os << tabs
            << formatv("for (; {0} < {1}.size(); ++{0})", wordIndex, words);

mlir/unittests/Interfaces/DataLayoutInterfacesTest.cpp

static unsigned getTypeSizeInBits(Type type, const DataLayout &dataLayout,
     // Make a recursive query.
     if (isa<FloatType>(type))
       return dataLayout.getTypeSizeInBits(
           IntegerType::get(type.getContext(), type.getIntOrFloatBitWidth()));
 
     // Handle built-in types that are not handled by the default process.
     if (auto iType = dyn_cast<IntegerType>(type)) {
       for (DataLayoutEntryInterface entry : params)
-        if (entry.getKey().dyn_cast<Type>() == type)
+        if (llvm::dyn_cast_if_present<Type>(entry.getKey()) == type)
           return 8 *
                  cast<IntegerAttr>(entry.getValue()).getValue().getZExtValue();
       return 8 * iType.getIntOrFloatBitWidth();
     }
 
     // Use the default process for everything else.
     return detail::getDefaultTypeSize(type, dataLayout, params);
   }