[SYCL][NFC] Fix some 'startswith/endswith' related to SYCL

clang/lib/CodeGen/CodeGenTypes.cpp

@@ -342,8 +342,8 @@ llvm::Type *CodeGenTypes::ConvertSYCLJointMatrixINTELType(RecordDecl *RD) {
  if (CompTy->isStructTy()) {
  StringRef LlvmTyName = CompTy->getStructName();
  // Emit half/int16/float for sycl[::*]::{half,bfloat16,tf32}
- if (LlvmTyName.startswith("class.sycl::") ||
- LlvmTyName.startswith("class.__sycl_internal::"))
+ if (LlvmTyName.starts_with("class.sycl::") ||
+ LlvmTyName.starts_with("class.__sycl_internal::"))
  LlvmTyName = LlvmTyName.rsplit("::").second;
  if (LlvmTyName == "half") {
  CompTy = llvm::Type::getHalfTy(getLLVMContext());

clang/lib/Driver/Driver.cpp

@@ -3518,7 +3518,7 @@ getLinkerArgs(Compilation &C, DerivedArgList &Args, bool IncludeObj = false) {
  // manner than the OpenMP processing. We should try and refactor this
  // to use the OpenMP flow (adding -l<name> to the llvm-link step)
  auto resolveStaticLib = [&](StringRef LibName, bool IsStatic) -> bool {
- if (!LibName.startswith("-l"))
+ if (!LibName.starts_with("-l"))
  return false;
  for (auto &LPath : LibPaths) {
  if (!IsStatic) {
@@ -3663,7 +3663,7 @@ static bool IsSYCLDeviceLibObj(std::string ObjFilePath, bool isMSVCEnv) {
  StringRef ObjFileName = llvm::sys::path::filename(ObjFilePath);
  StringRef ObjSuffix = isMSVCEnv ? ".obj" : ".o";
  bool Ret =
- (ObjFileName.startswith("libsycl-") && ObjFileName.endswith(ObjSuffix))
+ (ObjFileName.starts_with("libsycl-") && ObjFileName.ends_with(ObjSuffix))
  ? true
  : false;
  return Ret;

clang/lib/Driver/ToolChains/Linux.cpp

@@ -344,7 +344,7 @@ Linux::Linux(const Driver &D, const llvm::Triple &Triple, const ArgList &Args)
  // The deprecated -DLLVM_ENABLE_PROJECTS=libcxx configuration installs
  // libc++.so in D.Dir+"/../lib/". Detect this path.
  // TODO Remove once LLVM_ENABLE_PROJECTS=libcxx is unsupported.
- if (StringRef(D.Dir).startswith(SysRoot) &&
+ if (StringRef(D.Dir).starts_with(SysRoot) &&
  (Args.hasArg(options::OPT_fsycl) ||
  D.getVFS().exists(D.Dir + "/../lib/libsycl.so")))
  addPathIfExists(D, D.Dir + "/../lib", Paths);

clang/lib/Driver/ToolChains/SYCL.cpp

@@ -401,8 +401,9 @@ const char *SYCL::Linker::constructLLVMLinkCommand(
  LibPostfix = ".cubin";
  }
  StringRef LibSyclPrefix("libsycl-");
- if (!InputFilename.startswith(LibSyclPrefix) ||
- !InputFilename.endswith(LibPostfix) || (InputFilename.count('-') < 2))
+ if (!InputFilename.starts_with(LibSyclPrefix) ||
+ !InputFilename.ends_with(LibPostfix) ||
+ (InputFilename.count('-') < 2))
  return false;
  // Skip the prefix "libsycl-"
  std::string PureLibName =
@@ -419,7 +420,7 @@ const char *SYCL::Linker::constructLLVMLinkCommand(
  PureLibName.substr(0, FinalDashPos) + PureLibName.substr(DotPos);
  }
  for (const auto &L : SYCLDeviceLibList) {
- if (StringRef(PureLibName).startswith(L))
+ if (StringRef(PureLibName).starts_with(L))
  return true;
  }
  return false;
@@ -1354,7 +1355,7 @@ void SYCLToolChain::AddImpliedTargetArgs(const llvm::Triple &Triple,
  for (auto *A : Args) {
  if (!A->getOption().matches(options::OPT_Xsycl_backend_EQ))
  continue;
- if (StringRef(A->getValue()).startswith("intel_gpu"))
+ if (StringRef(A->getValue()).starts_with("intel_gpu"))
  TargArgs.push_back(A->getValue(1));
  }
  if (llvm::find_if(TargArgs, [&](auto Cur) {

clang/lib/Driver/ToolChains/SYCL.h

@@ -125,7 +125,7 @@ constexpr char AmdGPU[] = "amd_gpu_";
 template <auto GPUArh> std::optional<StringRef> isGPUTarget(StringRef Target) {
  // Handle target specifications that resemble '(intel, nvidia, amd)_gpu_*'
  // here.
- if (Target.startswith(GPUArh)) {
+ if (Target.starts_with(GPUArh)) {
  return resolveGenDevice(Target);
  }
  return std::nullopt;

clang/lib/Sema/SemaExpr.cpp

@@ -282,11 +282,11 @@ bool Sema::DiagnoseUseOfDecl(NamedDecl *D, ArrayRef<SourceLocation> Locs,
  .Default(false);
  };
  if ((getEmissionReason(FDecl) == Sema::DeviceDiagnosticReason::Sycl) &&
- Id && !Id->getName().startswith("__spirv_") &&
- !Id->getName().startswith("__sycl_") &&
- !Id->getName().startswith("__devicelib_ConvertBF16ToFINTEL") &&
- !Id->getName().startswith("__devicelib_ConvertFToBF16INTEL") &&
- !Id->getName().startswith("__assert_fail") &&
+ Id && !Id->getName().starts_with("__spirv_") &&
+ !Id->getName().starts_with("__sycl_") &&
+ !Id->getName().starts_with("__devicelib_ConvertBF16ToFINTEL") &&
+ !Id->getName().starts_with("__devicelib_ConvertFToBF16INTEL") &&
+ !Id->getName().starts_with("__assert_fail") &&
  !isMsvcMathFn(Id->getName())) {
  SYCLDiagIfDeviceCode(
  *Locs.begin(), diag::err_sycl_device_function_is_called_from_esimd,

clang/lib/Sema/SemaSYCL.cpp

@@ -4295,7 +4295,7 @@ static void CheckSYCL2020SubGroupSizes(Sema &S, FunctionDecl *SYCLKernel,
  // No need to validate __spirv routines here since they
  // are mapped to the equivalent SPIRV operations.
  const IdentifierInfo *II = FD->getIdentifier();
- if (II && II->getName().startswith("__spirv_"))
+ if (II && II->getName().starts_with("__spirv_"))
  return;
 
  // Else we need to figure out why they don't match.

clang/tools/clang-offload-deps/ClangOffloadDeps.cpp

@@ -173,7 +173,7 @@ int main(int argc, const char **argv) {
  // possibly reusing ClangOffloadBundler's 'OffloadTargetInfo'.
  for (const std::string &Target : Targets) {
  std::string Prefix = Target + ".";
- if (Symbol.startswith(Prefix))
+ if (Symbol.starts_with(Prefix))
  Target2Symbols[Target].insert(
  Symbol.substr(Prefix.size(), Len - Prefix.size()));
  }

llvm/lib/SYCLLowerIR/ESIMD/ESIMDUtils.cpp

@@ -34,13 +34,13 @@ constexpr char SLM_ALLOCATOR_DTOR_SUFFIX[] = "EED2Ev";
 bool isSlmAllocatorConstructor(const Function &F) {
  auto Name = F.getName();
  return Name.starts_with(SLM_ALLOCATOR_CTOR_DTOR_PREFIX) &&
- Name.endswith(SLM_ALLOCATOR_CTOR_SUFFIX);
+ Name.ends_with(SLM_ALLOCATOR_CTOR_SUFFIX);
 }
 
 bool isSlmAllocatorDestructor(const Function &F) {
  auto Name = F.getName();
  return Name.starts_with(SLM_ALLOCATOR_CTOR_DTOR_PREFIX) &&
- Name.endswith(SLM_ALLOCATOR_DTOR_SUFFIX);
+ Name.ends_with(SLM_ALLOCATOR_DTOR_SUFFIX);
 }
 
 bool isSlmInit(const Function &F) {

llvm/lib/SYCLLowerIR/ESIMD/LowerESIMD.cpp

@@ -1596,15 +1596,15 @@ SmallPtrSet<Type *, 4> collectGenXVolatileTypes(Module &M) {
  // TODO FIXME relying on type name in LLVM IR is fragile, needs rework
  if (!GTy || !GTy->getName()
  .rtrim(".0123456789")
- .endswith("sycl::_V1::ext::intel::esimd::simd"))
+ .ends_with("sycl::_V1::ext::intel::esimd::simd"))
  continue;
  assert(GTy->getNumContainedTypes() == 1);
  auto VTy = GTy->getContainedType(0);
  if ((GTy = dyn_cast<StructType>(VTy))) {
- assert(
- GTy->getName()
- .rtrim(".0123456789")
- .endswith("sycl::_V1::ext::intel::esimd::detail::simd_obj_impl"));
+ assert(GTy->getName()
+  .rtrim(".0123456789")
+  .ends_with(
+  "sycl::_V1::ext::intel::esimd::detail::simd_obj_impl"));
  VTy = GTy->getContainedType(0);
  }
  assert(VTy->isVectorTy());

llvm/lib/SYCLLowerIR/PrepareSYCLNativeCPU.cpp

@@ -283,7 +283,7 @@ static Function *addGetFunc(Module &M, StringRef Name, Type *StateType) {
 static Function *addReplaceFunc(Module &M, StringRef Name, Type *StateType) {
  Function *Res;
  const char GetPrefix[] = "__dpcpp_nativecpu_get";
- if (Name.startswith(GetPrefix)) {
+ if (Name.starts_with(GetPrefix)) {
  Res = addGetFunc(M, Name, StateType);
  } else if (Name == NativeCPUSetLocalId) {
  Res = addSetLocalIdFunc(M, Name, StateType);

llvm/lib/Support/PropertySetIO.cpp

@@ -37,7 +37,7 @@ PropertySetRegistry::read(const MemoryBuffer *Buf) {
 
  for (line_iterator LI(*Buf); !LI.is_at_end(); LI++) {
  // see if this line starts a new property set
- if (LI->startswith("[")) {
+ if (LI->starts_with("[")) {
  // yes - parse the category (property name)
  auto EndPos = LI->rfind(']');
  if (EndPos == StringRef::npos)

llvm/lib/Support/SimpleTable.cpp

@@ -214,8 +214,8 @@ Expected<SimpleTable::UPtrTy> SimpleTable::read(MemoryBuffer *Buf,
  return std::make_unique<SimpleTable>();
  UPtrTy Res;
 
- if (LI->startswith(COL_TITLE_LINE_OPEN)) {
- if (!LI->endswith(COL_TITLE_LINE_CLOSE))
+ if (LI->starts_with(COL_TITLE_LINE_OPEN)) {
+ if (!LI->ends_with(COL_TITLE_LINE_CLOSE))
  return createStringError(errc::invalid_argument, "malformed title line");
  // column titles present
  StringRef L = LI->substr(1, LI->size() - 2); // trim '[' and ']'

llvm/lib/TargetParser/Triple.cpp

@@ -726,7 +726,7 @@ static Triple::SubArchType parseSubArch(StringRef SubArchName) {
  (SubArchName.ends_with("r6el") || SubArchName.ends_with("r6")))
  return Triple::MipsSubArch_r6;
 
- if (SubArchName.startswith("spir")) {
+ if (SubArchName.starts_with("spir")) {
  StringRef SA(SubArchName);
  if (SA.consume_front("spir64_") || SA.consume_front("spir_")) {
  if (SA == "fpga")

llvm/lib/Transforms/InstCombine/InstCombineCasts.cpp

@@ -838,7 +838,7 @@ Instruction *InstCombinerImpl::visitTrunc(TruncInst &Trunc) {
  // // extractelement <8 x i32> (bitcast <4 x i64> %X to <8 x i32>), i32 0
  // ```
  // can't be lowered by SPIR-V translator to "standard" format.
- if (StringRef(Trunc.getModule()->getTargetTriple()).startswith("spir"))
+ if (StringRef(Trunc.getModule()->getTargetTriple()).starts_with("spir"))
  return nullptr;
 
  // Whenever an element is extracted from a vector, and then truncated,

llvm/lib/Transforms/Instrumentation/SPIRITTAnnotations.cpp

@@ -243,7 +243,7 @@ bool insertAtomicInstrumentationCall(Module &M, StringRef Name,
 
 PreservedAnalyses SPIRITTAnnotationsPass::run(Module &M,
  ModuleAnalysisManager &MAM) {
- assert(StringRef(M.getTargetTriple()).startswith("spir"));
+ assert(StringRef(M.getTargetTriple()).starts_with("spir"));
  bool IRModified = false;
  std::vector<StringRef> SPIRVCrossWGInstuctions = {
  SPIRV_CONTROL_BARRIER, SPIRV_GROUP_ALL, SPIRV_GROUP_ANY,
@@ -299,15 +299,15 @@ PreservedAnalyses SPIRITTAnnotationsPass::run(Module &M,
  if (std::any_of(SPIRVCrossWGInstuctions.begin(),
  SPIRVCrossWGInstuctions.end(),
  [&CalleeName](StringRef Name) {
- return CalleeName.startswith(Name);
+ return CalleeName.starts_with(Name);
  })) {
  Instruction *InstAfterBarrier = CI->getNextNode();
  const DebugLoc &DL = CI->getDebugLoc();
  insertSimpleInstrumentationCall(M, ITT_ANNOTATION_WG_BARRIER, CI, DL);
  insertSimpleInstrumentationCall(M, ITT_ANNOTATION_WI_RESUME,
  InstAfterBarrier, DL);
  IRModified = true;
- } else if (CalleeName.startswith(SPIRV_ATOMIC_INST)) {
+ } else if (CalleeName.starts_with(SPIRV_ATOMIC_INST)) {
  Instruction *InstAfterAtomic = CI->getNextNode();
  IRModified |= insertAtomicInstrumentationCall(
  M, ITT_ANNOTATION_ATOMIC_START, CI, CI, CalleeName);

llvm/tools/sycl-post-link/ModuleSplitter.cpp

@@ -79,7 +79,7 @@ bool isSpirvSyclBuiltin(StringRef FName) {
  // now skip the digits
  FName = FName.drop_while([](char C) { return std::isdigit(C); });
 
- return FName.startswith("__spirv_") || FName.startswith("__sycl_");
+ return FName.starts_with("__spirv_") || FName.starts_with("__sycl_");
 }
 
 // Return true if the function is a ESIMD builtin
@@ -91,12 +91,12 @@ bool isESIMDBuiltin(StringRef FName) {
  // now skip the digits
  FName = FName.drop_while([](char C) { return std::isdigit(C); });
 
- return FName.startswith("__esimd_");
+ return FName.starts_with("__esimd_");
 }
 
 // Return true if the function name starts with "__builtin_"
 bool isGenericBuiltin(StringRef FName) {
- return FName.startswith("__builtin_");
+ return FName.starts_with("__builtin_");
 }
 
 bool isKernel(const Function &F) {

llvm/tools/sycl-post-link/SYCLDeviceLibReqMask.cpp

@@ -696,7 +696,7 @@ uint32_t llvm::getSYCLDeviceLibReqMask(const Module &M) {
  return 0;
  uint32_t ReqMask = 0;
  for (const Function &SF : M) {
- if (SF.getName().startswith(DEVICELIB_FUNC_PREFIX) && SF.isDeclaration()) {
+ if (SF.getName().starts_with(DEVICELIB_FUNC_PREFIX) && SF.isDeclaration()) {
  assert(SF.getCallingConv() == CallingConv::SPIR_FUNC);
  uint32_t DeviceLibBits = getDeviceLibBits(SF.getName().str());
  ReqMask |= DeviceLibBits;

llvm/tools/sycl-post-link/SpecConstants.cpp

@@ -801,8 +801,8 @@ PreservedAnalyses SpecConstantsPass::run(Module &M,
  if (!F.isDeclaration())
  continue;
 
- if (!F.getName().startswith(SYCL_GET_SCALAR_2020_SPEC_CONST_VAL) &&
- !F.getName().startswith(SYCL_GET_COMPOSITE_2020_SPEC_CONST_VAL))
+ if (!F.getName().starts_with(SYCL_GET_SCALAR_2020_SPEC_CONST_VAL) &&
+ !F.getName().starts_with(SYCL_GET_COMPOSITE_2020_SPEC_CONST_VAL))
  continue;
 
  SmallVector<CallInst *, 32> SCIntrCalls;
@@ -1014,8 +1014,8 @@ bool SpecConstantsPass::collectSpecConstantDefaultValuesMetadata(
 
 bool llvm::checkModuleContainsSpecConsts(const Module &M) {
  for (const Function &F : M.functions()) {
- if (F.getName().startswith(SYCL_GET_SCALAR_2020_SPEC_CONST_VAL) ||
- F.getName().startswith(SYCL_GET_COMPOSITE_2020_SPEC_CONST_VAL))
+ if (F.getName().starts_with(SYCL_GET_SCALAR_2020_SPEC_CONST_VAL) ||
+ F.getName().starts_with(SYCL_GET_COMPOSITE_2020_SPEC_CONST_VAL))
  return true;
  }
 

llvm/tools/sycl-post-link/sycl-post-link.cpp

@@ -340,7 +340,7 @@ std::string makeResultFileName(Twine Ext, int I, StringRef Suffix) {
  : sys::path::parent_path(OutputFilename);
  const StringRef Sep = sys::path::get_separator();
  std::string Dir = Dir0.str();
- if (!Dir0.empty() && !Dir0.endswith(Sep))
+ if (!Dir0.empty() && !Dir0.ends_with(Sep))
  Dir += Sep.str();
  return Dir + sys::path::stem(OutputFilename).str() + Suffix.str() + "_" +
  std::to_string(I) + Ext.str();