[SYCL-MLIR][NFC] use cast<X>(Y) instead of Y.cast<X>() (#8754)

`Y.cast<X>()` is [deprecated](https://mlir.llvm.org/deprecation/).

---------

Signed-off-by: Tsang, Whitney <[email protected]>

Author: whitneywhtsang

mlir-sycl/lib/Conversion/SYCLToGPU/SYCLToGPU.cpp

@@ -75,7 +75,7 @@ Value createGetOp(OpBuilder &builder, Location loc, Type underlyingArrTy,
                   Value res, Value index, ArrayAttr argumentTypes,
                   FlatSymbolRefAttr functionName) {
   return TypeSwitch<Type, Value>(
-             res.getType().cast<MemRefType>().getElementType())
+             cast<MemRefType>(res.getType()).getElementType())
       .Case<IDType, RangeType>([&](auto arg) {
         // `this` type
         using ArgTy = decltype(arg);

mlir-sycl/lib/Conversion/SYCLToLLVM/SYCLFuncRegistry.cpp

@@ -37,7 +37,7 @@ Value SYCLFuncDescriptor::call(FuncId funcId, ValueRange args,
                    << funcDesc.funcRef << "\n");
 
   SmallVector<Type, 4> funcOutputTys;
-  if (!funcDesc.outputTy.isa<LLVM::LLVMVoidType>())
+  if (!isa<LLVM::LLVMVoidType>(funcDesc.outputTy))
     funcOutputTys.emplace_back(funcDesc.outputTy);
 
   LLVMBuilder builder(b, loc);

mlir-sycl/lib/Conversion/SYCLToLLVM/SYCLToLLVM.cpp

@@ -63,7 +63,7 @@ Value createGetOp(OpBuilder &builder, Location loc, Type dimMtTy, Value res,
                   Value index, ArrayAttr argumentTypes,
                   FlatSymbolRefAttr functionName) {
   return TypeSwitch<Type, Value>(
-             res.getType().cast<MemRefType>().getElementType())
+             cast<MemRefType>(res.getType()).getElementType())
       .Case<IDType, RangeType>([&](auto arg) {
         // `this` type
         using ArgTy = decltype(arg);

mlir-sycl/lib/Conversion/SYCLToSPIRV/SYCLToSPIRV.cpp

@@ -21,7 +21,7 @@ using namespace mlir;
 
 // Return true if the value \p val is a function argument, and false otherwise.
 static bool isFuncArg(Value val) {
-  auto blockArg = val.dyn_cast<BlockArgument>();
+  auto blockArg = dyn_cast<BlockArgument>(val);
   if (!blockArg)
     return false;
 
@@ -35,7 +35,7 @@ static bool isRestrict(Value val) {
   if (!isFuncArg(val))
     return false;
 
-  auto blockArg = val.cast<BlockArgument>();
+  auto blockArg = cast<BlockArgument>(val);
   auto func = cast<FunctionOpInterface>(blockArg.getOwner()->getParentOp());
   return !!func.getArgAttr(blockArg.getArgNumber(),
                            "local_alias_analysis.restrict");

mlir-sycl/lib/Dialect/Analysis/AliasAnalysis.cpp

@@ -20,8 +20,8 @@ bool SYCLCastOp::areCastCompatible(TypeRange Inputs, TypeRange Outputs) {
   if (Inputs.size() != 1 || Outputs.size() != 1)
     return false;
 
-  const auto Input = Inputs.front().dyn_cast<MemRefType>();
-  const auto Output = Outputs.front().dyn_cast<MemRefType>();
+  const auto Input = dyn_cast<MemRefType>(Inputs.front());
+  const auto Output = dyn_cast<MemRefType>(Outputs.front());
   if (!Input || !Output)
     return false;
 
@@ -69,8 +69,8 @@ bool SYCLAddrSpaceCastOp::areCastCompatible(TypeRange inputs,
   if (inputs.size() != 1 || outputs.size() != 1)
     return false;
 
-  const auto input = inputs.front().dyn_cast<MemRefType>();
-  const auto output = outputs.front().dyn_cast<MemRefType>();
+  const auto input = dyn_cast<MemRefType>(inputs.front());
+  const auto output = dyn_cast<MemRefType>(outputs.front());
   if (!input || !output)
     return false;
 
@@ -133,7 +133,7 @@ LogicalResult SYCLAccessorSubscriptOp::verify() {
             [&](auto Ty) { return VerifyElemType(Ty.getElementType()); })
         .Case<LLVM::LLVMPointerType>([&](auto Ty) {
           const Type ElemType = Ty.getElementType();
-          return (!ElemType.isa<LLVM::LLVMStructType>())
+          return (!isa<LLVM::LLVMStructType>(ElemType))
                      ? emitOpError(
                            "Expecting pointer to struct return type. Got ")
                            << ResultType

mlir-sycl/lib/Dialect/IR/SYCLOps.cpp

@@ -220,7 +220,7 @@ mlir::sycl::VecType::verify(llvm::function_ref<InFlightDiagnostic()> EmitError,
 }
 
 unsigned mlir::sycl::getDimensions(mlir::Type Type) {
-  if (auto MemRefTy = Type.dyn_cast<mlir::MemRefType>())
+  if (auto MemRefTy = dyn_cast<mlir::MemRefType>(Type))
     Type = MemRefTy.getElementType();
   return TypeSwitch<mlir::Type, unsigned>(Type)
       .Case<AccessorType, GroupType, IDType, ItemType, NdItemType, NdRangeType,

mlir-sycl/lib/Dialect/IR/SYCLOpsTypes.cpp

@@ -115,7 +115,7 @@ mlir::LogicalResult mlir::sycl::verifySYCLGetIDTrait(Operation *OpPtr) {
   const bool IsSizeTCast = FuncName == "operator unsigned long";
   const bool IsSubscript = FuncName == "operator[]";
   const mlir::Type RetTy = Op->getResult(0).getType();
-  const bool IsRetScalar = RetTy.isa<mlir::sycl::IDType>();
+  const bool IsRetScalar = isa<mlir::sycl::IDType>(RetTy);
   // operator size_t cannot be checked the generic way.
   if (FuncName != "operator unsigned long") {
     const LogicalResult GenericVerification =
@@ -171,7 +171,7 @@ mlir::LogicalResult mlir::sycl::verifySYCLGetGroupTrait(Operation *Op) {
 static LogicalResult verifyIndexSpaceTrait(Operation *Op) {
   const auto Ty = Op->getResultTypes();
   assert(Ty.size() == 1 && "Expecting a single return value");
-  const auto IsIndex = Ty[0].isa<IndexType>();
+  const auto IsIndex = isa<IndexType>(Ty[0]);
   switch (Op->getNumOperands()) {
   case 0:
     return !IsIndex ? success()

mlir-sycl/lib/Dialect/IR/SYCLTraits.cpp

@@ -30,8 +30,8 @@ static Value adaptArgumentForSYCLCall(OpBuilder &rewriter, Location loc,
   if (original.getType() == targetType)
     return original;
 
-  const auto mt = targetType.cast<MemRefType>();
-  const auto thisType = original.getType().cast<MemRefType>();
+  const auto mt = cast<MemRefType>(targetType);
+  const auto thisType = cast<MemRefType>(original.getType());
   const llvm::ArrayRef<int64_t> targetShape = mt.getShape();
   const Type targetElementType = mt.getElementType();
   const unsigned targetMemSpace = mt.getMemorySpaceAsInt();