[WIP] Normalize String Attributes

[serge-sans-paille]

Summary

LLVM IR supports free form string attributes. Those are used both internally in various passes, or to convey information from the front-end to the middle-end, and externally by third party projects.

This patch normalizes the internal part while still being open to extension for third-party project. The basic idea is to have one immutable object per string attribute, with its hash computed at compile-time. Then instead of referencing attributes per their name, one must reference them by their key.

Pros:

• avoid typos in attribute spelling etc, turn them in compile-time error
• it is possible to list all internal keys in one place, and eventually (not implemented in this patch) to normalize them (we currently have a mixture of underscores and hyphens). And to document them.
• hashes are precomputed, which leads to faster indexing and faster testing for attribute presence

[serge-sans-paille]

This is a WIP, to start the discussion. I've only turned on the X86 backend. Otherwise should be functional.

[Nuullll]

Thanks for this. I planned something similar to turn string literal typo errors into compile errors as well.
The typos are really hard to identify :-P

Some suggestions on using tablegen.

llvm/include/llvm/IR/Attributes.h

977–1133

I think we can simplify these definitions via "Attributes.inc", by adding one more class tablegen class and defining the symbolic -> display name mapping in "Attributes.td".

// llvm/utils/TableGen/Attributes.cpp emitTargetsIndependentNames()
Emit({"EnumAttr", "TypeAttr", "IntAttr"}, "ATTRIBUTE_ENUM");
Emit({"StrBoolAttr"}, "ATTRIBUTE_STRBOOL");
Emit({"StrKeyAttr"}, "ATTRIBUTE_KEY");

llvm/include/llvm/IR/Attributes.td

43

Define a dedicated tablegen class for AttributeKey generation.

303

Map symbolic names with string literals.

[mehdi_amini]

The patch is pretty large, any chance you could split in two? First the change introducing the AttributeKey and changing the API, and then updating all the callees separately as a follow up?

llvm/include/llvm/IR/Attributes.h

84

This whole code deserves documentation I think.

llvm/lib/IR/Attributes.cpp

125

Seems like this method does not use anything than the StringRef inside the Kind, why change it to take an AttributeKey? Won't you eagerly compute the hash even if not needed?

[serge-sans-paille]

This diff is indeed very large. I reworked it in https://reviews.llvm.org/D114394 to capture the essence of it, on which we could then rebase this patch.

[rnk]

it is possible to list all internal keys in one place

I think one of the original goals of adding support for string attributes (maybe @void will give some input) was specifically to avoid having one giant list with all the attributes supported by all the backends. String attributes were initially called "target dependent attributes", or something like that. I think we had a td_attrs accessor at some point. There was this idea that LLVM should not know about x86-specific attributes. There should be some kind of indirection and delegation to the target backend. LLVM has not always succeeded in following this vision, see for example the target-specific intrinsic enums, which were one giant enum until D71320.

I think, even if LLVM IR has to know all of the target-specific internally used attributes, we can try to honor that target independent vision by having separate headers for different target specific attributes. Does that sound reasonable? It's at least consistent with how we handle intrinsics (IntrinsicsX86.h, IntrinsicsAArch64.h etc).

[serge-sans-paille]

In D114082#3149631, @rnk wrote:

it is possible to list all internal keys in one place

I think one of the original goals of adding support for string attributes (maybe @void will give some input) was specifically to avoid having one giant list with all the attributes supported by all the backends. String attributes were initially called "target dependent attributes", or something like that. I think we had a td_attrs accessor at some point. There was this idea that LLVM should not know about x86-specific attributes. There should be some kind of indirection and delegation to the target backend. LLVM has not always succeeded in following this vision, see for example the target-specific intrinsic enums, which were one giant enum until D71320.

I think, even if LLVM IR has to know all of the target-specific internally used attributes, we can try to honor that target independent vision by having separate headers for different target specific attributes. Does that sound reasonable? It's at least consistent with how we handle intrinsics (IntrinsicsX86.h, IntrinsicsAArch64.h etc).

The sub-review I proposed in https://reviews.llvm.org/D114394 does not introduce any list, so that should be a decent first step. As a second step, many passes already register their attribute in some local header and I can just generalize that: no big list but some local explicit definition.

clang/lib/CodeGen/CGBlocks.cpp

   unsigned AddrSpace = 0;
   if (CGM.getContext().getLangOpts().OpenCL)
     AddrSpace = CGM.getContext().getTargetAddressSpace(LangAS::opencl_global);
 
   llvm::GlobalVariable *literal = fields.finishAndCreateGlobal(
       "__block_literal_global", blockInfo.BlockAlign,
       /*constant*/ !IsWindows, llvm::GlobalVariable::InternalLinkage, AddrSpace);
 
-  literal->addAttribute("objc_arc_inert");
+  literal->addAttribute(llvm::ObjcArcInertAttr);
 
   // Windows does not allow globals to be initialised to point to globals in
   // different DLLs.  Any such variables must run code to initialise them.
   if (IsWindows) {
     auto *Init = llvm::Function::Create(llvm::FunctionType::get(CGM.VoidTy,
           {}), llvm::GlobalValue::InternalLinkage, ".block_isa_init",
         &CGM.getModule());
     llvm::IRBuilder<> b(llvm::BasicBlock::Create(CGM.getLLVMContext(), "entry",

clang/lib/CodeGen/CGCall.cpp

     return;
 
   SmallVector<StringRef, 4> Attrs;
 
   for (const AssumptionAttr *AA : Callee->specific_attrs<AssumptionAttr>())
     AA->getAssumption().split(Attrs, ",");
 
   if (!Attrs.empty())
-    FuncAttrs.addAttribute(llvm::AssumptionAttrKey,
+    FuncAttrs.addAttribute(llvm::LLVMAssumeAttr,
                            llvm::join(Attrs.begin(), Attrs.end(), ","));
 }
 
 bool CodeGenModule::MayDropFunctionReturn(const ASTContext &Context,
                                           QualType ReturnType) {
   // We can't just discard the return value for a record type with a
   // complex destructor or a non-trivially copyable type.
   if (const RecordType *RT =

     if (CodeGenOpts.OptimizeSize)
       FuncAttrs.addAttribute(llvm::Attribute::OptimizeForSize);
     if (CodeGenOpts.OptimizeSize == 2)
       FuncAttrs.addAttribute(llvm::Attribute::MinSize);
   }
 
   if (CodeGenOpts.DisableRedZone)
     FuncAttrs.addAttribute(llvm::Attribute::NoRedZone);
-  if (CodeGenOpts.IndirectTlsSegRefs)
-    FuncAttrs.addAttribute("indirect-tls-seg-refs");
+  if (CodeGenOpts.IndirectTlsSegRefs) {
+    FuncAttrs.addAttribute(llvm::IndirectTLSSegRefsAttr);
+  }
   if (CodeGenOpts.NoImplicitFloat)
     FuncAttrs.addAttribute(llvm::Attribute::NoImplicitFloat);
 
   if (AttrOnCallSite) {
     // Attributes that should go on the call site only.
     if (!CodeGenOpts.SimplifyLibCalls || LangOpts.isNoBuiltinFunc(Name))
       FuncAttrs.addAttribute(llvm::Attribute::NoBuiltin);
-    if (!CodeGenOpts.TrapFuncName.empty())
-      FuncAttrs.addAttribute("trap-func-name", CodeGenOpts.TrapFuncName);
+    if (!CodeGenOpts.TrapFuncName.empty()) {
+      FuncAttrs.addAttribute(llvm::TrapFuncNameAttr, CodeGenOpts.TrapFuncName);
+    }
   } else {
     StringRef FpKind;
     switch (CodeGenOpts.getFramePointer()) {
     case CodeGenOptions::FramePointerKind::None:
       FpKind = "none";
       break;
     case CodeGenOptions::FramePointerKind::NonLeaf:
       FpKind = "non-leaf";
       break;
     case CodeGenOptions::FramePointerKind::All:
       FpKind = "all";
       break;
     }
-    FuncAttrs.addAttribute("frame-pointer", FpKind);
+    FuncAttrs.addAttribute(llvm::FramePointerAttr, FpKind);
 
-    if (CodeGenOpts.LessPreciseFPMAD)
-      FuncAttrs.addAttribute("less-precise-fpmad", "true");
+    if (CodeGenOpts.LessPreciseFPMAD) {
+      FuncAttrs.addAttribute(llvm::LessPreciseFPMADAttr, "true");
+    }
 
     if (CodeGenOpts.NullPointerIsValid)
       FuncAttrs.addAttribute(llvm::Attribute::NullPointerIsValid);
 
-    if (CodeGenOpts.FPDenormalMode != llvm::DenormalMode::getIEEE())
-      FuncAttrs.addAttribute("denormal-fp-math",
+    if (CodeGenOpts.FPDenormalMode != llvm::DenormalMode::getIEEE()) {
+      FuncAttrs.addAttribute(llvm::DenormalFPMathAttr,
                              CodeGenOpts.FPDenormalMode.str());
+    }
     if (CodeGenOpts.FP32DenormalMode != CodeGenOpts.FPDenormalMode) {
-      FuncAttrs.addAttribute(
-          "denormal-fp-math-f32",
-          CodeGenOpts.FP32DenormalMode.str());
+      FuncAttrs.addAttribute(llvm::DenormalFPMathf32Attr,
+                             CodeGenOpts.FP32DenormalMode.str());
     }
 
-    if (LangOpts.getFPExceptionMode() == LangOptions::FPE_Ignore)
-      FuncAttrs.addAttribute("no-trapping-math", "true");
+    if (LangOpts.getFPExceptionMode() == LangOptions::FPE_Ignore) {
+      FuncAttrs.addAttribute(llvm::NoTrappingMathAttr, "true");
+    }
 
     // Strict (compliant) code is the default, so only add this attribute to
     // indicate that we are trying to workaround a problem case.
-    if (!CodeGenOpts.StrictFloatCastOverflow)
-      FuncAttrs.addAttribute("strict-float-cast-overflow", "false");
+    if (!CodeGenOpts.StrictFloatCastOverflow) {
+      FuncAttrs.addAttribute(llvm::StrictFloatCastOverflowAttr, "false");
+    }
 
     // TODO: Are these all needed?
     // unsafe/inf/nan/nsz are handled by instruction-level FastMathFlags.
-    if (LangOpts.NoHonorInfs)
-      FuncAttrs.addAttribute("no-infs-fp-math", "true");
-    if (LangOpts.NoHonorNaNs)
-      FuncAttrs.addAttribute("no-nans-fp-math", "true");
-    if (LangOpts.ApproxFunc)
-      FuncAttrs.addAttribute("approx-func-fp-math", "true");
-    if (LangOpts.UnsafeFPMath)
-      FuncAttrs.addAttribute("unsafe-fp-math", "true");
-    if (CodeGenOpts.SoftFloat)
-      FuncAttrs.addAttribute("use-soft-float", "true");
-    FuncAttrs.addAttribute("stack-protector-buffer-size",
+    if (LangOpts.NoHonorInfs) {
+      FuncAttrs.addAttribute(llvm::NoInfsFPMathAttr, "true");
+    }
+    if (LangOpts.NoHonorNaNs) {
+      FuncAttrs.addAttribute(llvm::NoNansFPMathAttr, "true");
+    }
+    if (LangOpts.ApproxFunc) {
+      FuncAttrs.addAttribute(llvm::ApproxFuncFPMathAttr, "true");
+    }
+    if (LangOpts.UnsafeFPMath) {
+      FuncAttrs.addAttribute(llvm::UnsafeFPMathAttr, "true");
+    }
+    if (CodeGenOpts.SoftFloat) {
+      FuncAttrs.addAttribute(llvm::UseSoftFloatAttr, "true");
+    }
+    FuncAttrs.addAttribute(llvm::StackProtectorBufferSizeAttr,
                            llvm::utostr(CodeGenOpts.SSPBufferSize));
-    if (LangOpts.NoSignedZero)
-      FuncAttrs.addAttribute("no-signed-zeros-fp-math", "true");
+    if (LangOpts.NoSignedZero) {
+      FuncAttrs.addAttribute(llvm::NoSignedZerosFPMathAttr, "true");
+    }
 
     // TODO: Reciprocal estimate codegen options should apply to instructions?
     const std::vector<std::string> &Recips = CodeGenOpts.Reciprocals;
-    if (!Recips.empty())
-      FuncAttrs.addAttribute("reciprocal-estimates",
+    if (!Recips.empty()) {
+      FuncAttrs.addAttribute(llvm::ReciprocalEstimatesAttr,
                              llvm::join(Recips, ","));
+    }
 
     if (!CodeGenOpts.PreferVectorWidth.empty() &&
-        CodeGenOpts.PreferVectorWidth != "none")
-      FuncAttrs.addAttribute("prefer-vector-width",
+        CodeGenOpts.PreferVectorWidth != "none") {
+      FuncAttrs.addAttribute(llvm::PreferVectorWidthAttr,
                              CodeGenOpts.PreferVectorWidth);
+    }
 
-    if (CodeGenOpts.StackRealignment)
-      FuncAttrs.addAttribute("stackrealign");
-    if (CodeGenOpts.Backchain)
-      FuncAttrs.addAttribute("backchain");
-    if (CodeGenOpts.EnableSegmentedStacks)
-      FuncAttrs.addAttribute("split-stack");
+    if (CodeGenOpts.StackRealignment) {
+      FuncAttrs.addAttribute(llvm::StackrealignAttr);
+    }
+    if (CodeGenOpts.Backchain) {
+      FuncAttrs.addAttribute(llvm::BackchainAttr);
+    }
+    if (CodeGenOpts.EnableSegmentedStacks) {
+      FuncAttrs.addAttribute(llvm::SplitStackAttr);
+    }
 
     if (CodeGenOpts.SpeculativeLoadHardening)
       FuncAttrs.addAttribute(llvm::Attribute::SpeculativeLoadHardening);
   }
 
   if (getLangOpts().assumeFunctionsAreConvergent()) {
     // Conservatively, mark all functions and calls in CUDA and OpenCL as
     // convergent (meaning, they may call an intrinsically convergent op, such
     // as __syncthreads() / barrier(), and so can't have certain optimizations
     // applied around them).  LLVM will remove this attribute where it safely
     // can.
     FuncAttrs.addAttribute(llvm::Attribute::Convergent);
   }
 
   if (getLangOpts().CUDA && getLangOpts().CUDAIsDevice) {
     // Exceptions aren't supported in CUDA device code.
     FuncAttrs.addAttribute(llvm::Attribute::NoUnwind);
   }
 
   for (StringRef Attr : CodeGenOpts.DefaultFunctionAttrs) {
     StringRef Var, Value;
     std::tie(Var, Value) = Attr.split('=');
-    FuncAttrs.addAttribute(Var, Value);
+    FuncAttrs.addAttribute(llvm::AttributeKey::Create(Var), Value);
   }
 }
 
 void CodeGenModule::addDefaultFunctionDefinitionAttributes(llvm::Function &F) {
   llvm::AttrBuilder FuncAttrs;
   getDefaultFunctionAttributes(F.getName(), F.hasOptNone(),
                                /* AttrOnCallSite = */ false, FuncAttrs);
   // TODO: call GetCPUAndFeaturesAttributes?

 
 static void addNoBuiltinAttributes(llvm::AttrBuilder &FuncAttrs,
                                    const LangOptions &LangOpts,
                                    const NoBuiltinAttr *NBA = nullptr) {
   auto AddNoBuiltinAttr = [&FuncAttrs](StringRef BuiltinName) {
     SmallString<32> AttributeName;
     AttributeName += "no-builtin-";
     AttributeName += BuiltinName;
-    FuncAttrs.addAttribute(AttributeName);
+    FuncAttrs.addAttribute(llvm::AttributeKey::Create(AttributeName));
   };
 
   // First, handle the language options passed through -fno-builtin.
   if (LangOpts.NoBuiltin) {
     // -fno-builtin disables them all.
-    FuncAttrs.addAttribute("no-builtins");
+    FuncAttrs.addAttribute(llvm::NoBuiltinsAttr);
     return;
   }
 
   // Then, add attributes for builtins specified through -fno-builtin-<name>.
   llvm::for_each(LangOpts.NoBuiltinFuncs, AddNoBuiltinAttr);
 
   // Now, let's check the __attribute__((no_builtin("...")) attribute added to
   // the source.
   if (!NBA)
     return;
 
   // If there is a wildcard in the builtin names specified through the
   // attribute, disable them all.
   if (llvm::is_contained(NBA->builtinNames(), "*")) {
-    FuncAttrs.addAttribute("no-builtins");
+    FuncAttrs.addAttribute(llvm::NoBuiltinsAttr);
     return;
   }
 
   // And last, add the rest of the builtin names.
   llvm::for_each(NBA->builtinNames(), AddNoBuiltinAttr);
 }
 
 static bool DetermineNoUndef(QualType QTy, CodeGenTypes &Types,

   llvm::AttrBuilder FuncAttrs;
   llvm::AttrBuilder RetAttrs;
 
   // Collect function IR attributes from the CC lowering.
   // We'll collect the paramete and result attributes later.
   CallingConv = FI.getEffectiveCallingConvention();
   if (FI.isNoReturn())
     FuncAttrs.addAttribute(llvm::Attribute::NoReturn);
-  if (FI.isCmseNSCall())
-    FuncAttrs.addAttribute("cmse_nonsecure_call");
+  if (FI.isCmseNSCall()) {
+    FuncAttrs.addAttribute(llvm::CmseNonsecureEntryAttr);
+  }
 
   // Collect function IR attributes from the callee prototype if we have one.
   AddAttributesFromFunctionProtoType(getContext(), FuncAttrs,
                                      CalleeInfo.getCalleeFunctionProtoType());
 
   const Decl *TargetDecl = CalleeInfo.getCalleeDecl().getDecl();
 
   // Attach assumption attributes to the declaration. If this is a call

       FuncAttrs.addAttribute(llvm::Attribute::ArgMemOnly);
       FuncAttrs.addAttribute(llvm::Attribute::NoUnwind);
     }
     if (TargetDecl->hasAttr<RestrictAttr>())
       RetAttrs.addAttribute(llvm::Attribute::NoAlias);
     if (TargetDecl->hasAttr<ReturnsNonNullAttr>() &&
         !CodeGenOpts.NullPointerIsValid)
       RetAttrs.addAttribute(llvm::Attribute::NonNull);
-    if (TargetDecl->hasAttr<AnyX86NoCallerSavedRegistersAttr>())
-      FuncAttrs.addAttribute("no_caller_saved_registers");
+    if (TargetDecl->hasAttr<AnyX86NoCallerSavedRegistersAttr>()) {
+      FuncAttrs.addAttribute(llvm::NoCallerSavedRegistersAttr);
+    }
     if (TargetDecl->hasAttr<AnyX86NoCfCheckAttr>())
       FuncAttrs.addAttribute(llvm::Attribute::NoCfCheck);
     if (TargetDecl->hasAttr<LeafAttr>())
       FuncAttrs.addAttribute(llvm::Attribute::NoCallback);
 
     HasOptnone = TargetDecl->hasAttr<OptimizeNoneAttr>();
     if (auto *AllocSize = TargetDecl->getAttr<AllocSizeAttr>()) {
       Optional<unsigned> NumElemsParam;
       if (AllocSize->getNumElemsParam().isValid())
         NumElemsParam = AllocSize->getNumElemsParam().getLLVMIndex();
       FuncAttrs.addAllocSizeAttr(AllocSize->getElemSizeParam().getLLVMIndex(),
                                  NumElemsParam);
     }
 
     if (TargetDecl->hasAttr<OpenCLKernelAttr>()) {
       if (getLangOpts().OpenCLVersion <= 120) {
         // OpenCL v1.2 Work groups are always uniform
-        FuncAttrs.addAttribute("uniform-work-group-size", "true");
+        FuncAttrs.addAttribute(llvm::UniformWorkGroupSizeAttr, "true");
       } else {
         // OpenCL v2.0 Work groups may be whether uniform or not.
         // '-cl-uniform-work-group-size' compile option gets a hint
         // to the compiler that the global work-size be a multiple of
         // the work-group size specified to clEnqueueNDRangeKernel
         // (i.e. work groups are uniform).
-        FuncAttrs.addAttribute("uniform-work-group-size",
+        FuncAttrs.addAttribute(llvm::UniformWorkGroupSizeAttr,
                                llvm::toStringRef(CodeGenOpts.UniformWGSize));
       }
     }
   }
 
   // Attach "no-builtins" attributes to:
   // * call sites: both `nobuiltin` and "no-builtins" or "no-builtin-<name>".
   // * definitions: "no-builtins" or "no-builtin-<name>" only.
   // The attributes can come from:
   // * LangOpts: -ffreestanding, -fno-builtin, -fno-builtin-<name>
   // * FunctionDecl attributes: __attribute__((no_builtin(...)))
   addNoBuiltinAttributes(FuncAttrs, getLangOpts(), NBA);
 
   // Collect function IR attributes based on global settiings.
   getDefaultFunctionAttributes(Name, HasOptnone, AttrOnCallSite, FuncAttrs);
 
   // Override some default IR attributes based on declaration-specific
   // information.
   if (TargetDecl) {
     if (TargetDecl->hasAttr<NoSpeculativeLoadHardeningAttr>())
       FuncAttrs.removeAttribute(llvm::Attribute::SpeculativeLoadHardening);
     if (TargetDecl->hasAttr<SpeculativeLoadHardeningAttr>())
       FuncAttrs.addAttribute(llvm::Attribute::SpeculativeLoadHardening);
-    if (TargetDecl->hasAttr<NoSplitStackAttr>())
-      FuncAttrs.removeAttribute("split-stack");
+    if (TargetDecl->hasAttr<NoSplitStackAttr>()) {
+      FuncAttrs.removeAttribute(llvm::SplitStackAttr);
+    }
 
     // Add NonLazyBind attribute to function declarations when -fno-plt
     // is used.
     // FIXME: what if we just haven't processed the function definition
     // yet, or if it's an external definition like C99 inline?
     if (CodeGenOpts.NoPLT) {
       if (auto *Fn = dyn_cast<FunctionDecl>(TargetDecl)) {
         if (!Fn->isDefined() && !AttrOnCallSite) {
           FuncAttrs.addAttribute(llvm::Attribute::NonLazyBind);
         }
       }
     }
   }
 
   // Add "sample-profile-suffix-elision-policy" attribute for internal linkage
   // functions with -funique-internal-linkage-names.
   if (TargetDecl && CodeGenOpts.UniqueInternalLinkageNames) {
     if (isa<FunctionDecl>(TargetDecl)) {
       if (this->getFunctionLinkage(CalleeInfo.getCalleeDecl()) ==
-          llvm::GlobalValue::InternalLinkage)
-        FuncAttrs.addAttribute("sample-profile-suffix-elision-policy",
+          llvm::GlobalValue::InternalLinkage) {
+        FuncAttrs.addAttribute(llvm::SampleProfileSuffixElisionPolicyAttr,
                                "selected");
+      }
     }
   }
 
   // Collect non-call-site function IR attributes from declaration-specific
   // information.
   if (!AttrOnCallSite) {
-    if (TargetDecl && TargetDecl->hasAttr<CmseNSEntryAttr>())
-      FuncAttrs.addAttribute("cmse_nonsecure_entry");
+    if (TargetDecl && TargetDecl->hasAttr<CmseNSEntryAttr>()) {
+      FuncAttrs.addAttribute(llvm::CmseNonsecureEntryAttr);
+    }
 
     // Whether tail calls are enabled.
     auto shouldDisableTailCalls = [&] {
       // Should this be honored in getDefaultFunctionAttributes?
       if (CodeGenOpts.DisableTailCalls)
         return true;
 
       if (!TargetDecl)
         return false;
 
       if (TargetDecl->hasAttr<DisableTailCallsAttr>() ||
           TargetDecl->hasAttr<AnyX86InterruptAttr>())
         return true;
 
       if (CodeGenOpts.NoEscapingBlockTailCalls) {
         if (const auto *BD = dyn_cast<BlockDecl>(TargetDecl))
           if (!BD->doesNotEscape())
             return true;
       }
 
       return false;
     };
-    if (shouldDisableTailCalls())
-      FuncAttrs.addAttribute("disable-tail-calls", "true");
+    if (shouldDisableTailCalls()) {
+      FuncAttrs.addAttribute(llvm::DisableTailCallsAttr, "true");
+    }
 
     // CPU/feature overrides.  addDefaultFunctionDefinitionAttributes
     // handles these separately to set them based on the global defaults.
     GetCPUAndFeaturesAttributes(CalleeInfo.getCalleeDecl(), FuncAttrs);
   }
 
   // Collect attributes from arguments and return values.
   ClangToLLVMArgMapping IRFunctionArgs(getContext(), FI);

   if (callOrInvoke)
     *callOrInvoke = CI;
 
   // If this is within a function that has the guard(nocf) attribute and is an
   // indirect call, add the "guard_nocf" attribute to this call to indicate that
   // Control Flow Guard checks should not be added, even if the call is inlined.
   if (const auto *FD = dyn_cast_or_null<FunctionDecl>(CurFuncDecl)) {
     if (const auto *A = FD->getAttr<CFGuardAttr>()) {
-      if (A->getGuard() == CFGuardAttr::GuardArg::nocf && !CI->getCalledFunction())
-        Attrs = Attrs.addFnAttribute(getLLVMContext(), "guard_nocf");
+      if (A->getGuard() == CFGuardAttr::GuardArg::nocf &&
+          !CI->getCalledFunction()) {
+        Attrs = Attrs.addFnAttribute(getLLVMContext(), llvm::GuardNocfAttr);
+      }
     }
   }
 
   // Apply the attributes and calling convention.
   CI->setAttributes(Attrs);
   CI->setCallingConv(static_cast<llvm::CallingConv::ID>(CallingConv));
 
   // Apply various metadata.

clang/lib/CodeGen/CGClass.cpp

       const ASTRecordLayout &Layout =
           CGF.getContext().getASTRecordLayout(Dtor->getParent());
 
       // Nothing to poison.
       if (Layout.getFieldCount() == 0)
         return;
 
       // Prevent the current stack frame from disappearing from the stack trace.
-      CGF.CurFn->addFnAttr("disable-tail-calls", "true");
+      CGF.CurFn->addFnAttr(llvm::DisableTailCallsAttr, "true");
 
       // Construct pointer to region to begin poisoning, and calculate poison
       // size, so that only members declared in this class are poisoned.
       ASTContext &Context = CGF.getContext();
 
       const RecordDecl *Decl = Dtor->getParent();
       auto Fields = Decl->fields();
       auto IsTrivial = [&](const FieldDecl *F) {

clang/lib/CodeGen/CGDecl.cpp

void CodeGenFunction::EmitStaticVarDecl(const VarDecl &D,
   if (D.getInit() && !isCudaSharedVar)
     var = AddInitializerToStaticVarDecl(D, var);
 
   var->setAlignment(alignment.getAsAlign());
 
   if (D.hasAttr<AnnotateAttr>())
     CGM.AddGlobalAnnotations(&D, var);
 
-  if (auto *SA = D.getAttr<PragmaClangBSSSectionAttr>())
-    var->addAttribute("bss-section", SA->getName());
-  if (auto *SA = D.getAttr<PragmaClangDataSectionAttr>())
-    var->addAttribute("data-section", SA->getName());
-  if (auto *SA = D.getAttr<PragmaClangRodataSectionAttr>())
-    var->addAttribute("rodata-section", SA->getName());
-  if (auto *SA = D.getAttr<PragmaClangRelroSectionAttr>())
-    var->addAttribute("relro-section", SA->getName());
+  if (auto *SA = D.getAttr<PragmaClangBSSSectionAttr>()) {
+    var->addAttribute(llvm::BssSectionAttr, SA->getName());
+  }
+  if (auto *SA = D.getAttr<PragmaClangDataSectionAttr>()) {
+    var->addAttribute(llvm::DataSectionAttr, SA->getName());
+  }
+  if (auto *SA = D.getAttr<PragmaClangRodataSectionAttr>()) {
+    var->addAttribute(llvm::RodataSectionAttr, SA->getName());
+  }
+  if (auto *SA = D.getAttr<PragmaClangRelroSectionAttr>()) {
+    var->addAttribute(llvm::RelroSectionAttr, SA->getName());
+  }
 
   if (const SectionAttr *SA = D.getAttr<SectionAttr>())
     var->setSection(SA->getName());
 
   if (D.hasAttr<RetainAttr>())
     CGM.addUsedGlobal(var);
   else if (D.hasAttr<UsedAttr>())
     CGM.addUsedOrCompilerUsedGlobal(var);

clang/lib/CodeGen/CGDeclCXX.cpp

if (getLangOpts().OpenCL) {
     GenOpenCLArgMetadata(Fn);
     Fn->setCallingConv(llvm::CallingConv::SPIR_KERNEL);
   }
 
   assert(!getLangOpts().CUDA || !getLangOpts().CUDAIsDevice ||
          getLangOpts().GPUAllowDeviceInit);
   if (getLangOpts().HIP && getLangOpts().CUDAIsDevice) {
     Fn->setCallingConv(llvm::CallingConv::AMDGPU_KERNEL);
-    Fn->addFnAttr("device-init");
+    Fn->addFnAttr(llvm::DeviceInitAttr);
   }
 
   CXXGlobalInits.clear();
 }
 
 void CodeGenModule::EmitCXXGlobalCleanUpFunc() {
   if (CXXGlobalDtorsOrStermFinalizers.empty() &&
       PrioritizedCXXStermFinalizers.empty())

clang/lib/CodeGen/CGExpr.cpp

     Builder.CreateCondBr(Checked, Cont, TrapBB);
     EmitBlock(TrapBB);
 
     llvm::CallInst *TrapCall =
         Builder.CreateCall(CGM.getIntrinsic(llvm::Intrinsic::ubsantrap),
                            llvm::ConstantInt::get(CGM.Int8Ty, CheckHandlerID));
 
     if (!CGM.getCodeGenOpts().TrapFuncName.empty()) {
-      auto A = llvm::Attribute::get(getLLVMContext(), "trap-func-name",
+      auto A = llvm::Attribute::get(getLLVMContext(), llvm::TrapFuncNameAttr,
                                     CGM.getCodeGenOpts().TrapFuncName);
       TrapCall->addFnAttr(A);
     }
     TrapCall->setDoesNotReturn();
     TrapCall->setDoesNotThrow();
     Builder.CreateUnreachable();
   } else {
     auto Call = TrapBB->begin();
     assert(isa<llvm::CallInst>(Call) && "Expected call in trap BB");
 
     Call->applyMergedLocation(Call->getDebugLoc(),
                               Builder.getCurrentDebugLocation());
     Builder.CreateCondBr(Checked, Cont, TrapBB);
   }
 
   EmitBlock(Cont);
 }
 
 llvm::CallInst *CodeGenFunction::EmitTrapCall(llvm::Intrinsic::ID IntrID) {
   llvm::CallInst *TrapCall =
       Builder.CreateCall(CGM.getIntrinsic(IntrID));
 
   if (!CGM.getCodeGenOpts().TrapFuncName.empty()) {
-    auto A = llvm::Attribute::get(getLLVMContext(), "trap-func-name",
+    auto A = llvm::Attribute::get(getLLVMContext(), llvm::TrapFuncNameAttr,
                                   CGM.getCodeGenOpts().TrapFuncName);
     TrapCall->addFnAttr(A);
   }
 
   return TrapCall;
 }
 
 Address CodeGenFunction::EmitArrayToPointerDecay(const Expr *E,

clang/lib/CodeGen/CGOpenMPRuntime.cpp

   OffloadEntriesInfoManager.registerTargetRegionEntryInfo(
       DeviceID, FileID, ParentName, Line, OutlinedFn, OutlinedFnID,
       OffloadEntriesInfoManagerTy::OMPTargetRegionEntryTargetRegion);
 
   // Add NumTeams and ThreadLimit attributes to the outlined GPU function
   int32_t DefaultValTeams = -1;
   getNumTeamsExprForTargetDirective(CGF, D, DefaultValTeams);
   if (DefaultValTeams > 0) {
-    OutlinedFn->addFnAttr("omp_target_num_teams",
+    OutlinedFn->addFnAttr(llvm::OmpTargetNumTeamsAttr,
                           std::to_string(DefaultValTeams));
   }
   int32_t DefaultValThreads = -1;
   getNumThreadsExprForTargetDirective(CGF, D, DefaultValThreads);
   if (DefaultValThreads > 0) {
-    OutlinedFn->addFnAttr("omp_target_thread_limit",
+    OutlinedFn->addFnAttr(llvm::OmpTargetThreadLimitAttr,
                           std::to_string(DefaultValThreads));
   }
 }
 
 /// Checks if the expression is constant or does not have non-trivial function
 /// calls.
 static bool isTrivial(ASTContext &Ctx, const Expr * E) {
   // We can skip constant expressions.

         case Vector:
           Out << 'v';
           break;
         }
         if (!!ParamAttr.Alignment)
           Out << 'a' << ParamAttr.Alignment;
       }
       Out << '_' << Fn->getName();
-      Fn->addFnAttr(Out.str());
+      Fn->addFnAttr(llvm::AttributeKey::Create(Out.str()));
     }
   }
 }
 
 // This are the Functions that are needed to mangle the name of the
 // vector functions generated by the compiler, according to the rules
 // defined in the "Vector Function ABI specifications for AArch64",
 // available at

                                  StringRef MangledName, bool OutputBecomesInput,
                                  llvm::Function *Fn) {
   SmallString<256> Buffer;
   llvm::raw_svector_ostream Out(Buffer);
   Out << Prefix << ISA << LMask << VLEN;
   if (OutputBecomesInput)
     Out << "v";
   Out << ParSeq << "_" << MangledName;
-  Fn->addFnAttr(Out.str());
+  Fn->addFnAttr(llvm::AttributeKey::Create(Out.str()));
 }
 
 // Helper function to generate the Advanced SIMD names depending on
 // the value of the NDS when simdlen is not present.
 static void addAArch64AdvSIMDNDSNames(unsigned NDS, StringRef Mask,
                                       StringRef Prefix, char ISA,
                                       StringRef ParSeq, StringRef MangledName,
                                       bool OutputBecomesInput,

clang/lib/CodeGen/CGVTables.cpp

if (!ThunkFn->isDeclaration()) {
     setThunkProperties(CGM, TI, ThunkFn, ForVTable, GD);
     return ThunkFn;
   }
 
   // If this will be unprototyped, add the "thunk" attribute so that LLVM knows
   // that the return type is meaningless. These thunks can be used to call
   // functions with differing return types, and the caller is required to cast
   // the prototype appropriately to extract the correct value.
-  if (IsUnprototyped)
-    ThunkFn->addFnAttr("thunk");
+  if (IsUnprototyped) {
+    ThunkFn->addFnAttr(llvm::ThunkAttr);
+  }
 
   CGM.SetLLVMFunctionAttributesForDefinition(GD.getDecl(), ThunkFn);
 
   // Thunks for variadic methods are special because in general variadic
   // arguments cannot be perfectly forwarded. In the general case, clang
   // implements such thunks by cloning the original function body. However, for
   // thunks with no return adjustment on targets that support musttail, we can
   // use musttail to perfectly forward the variadic arguments.

clang/lib/CodeGen/CodeGenFunction.cpp

void CodeGenFunction::CGFPOptionsRAII::ConstructorHelper(FPOptions FPFeatures) {
 
   assert((CGF.CurFuncDecl == nullptr || CGF.Builder.getIsFPConstrained() ||
           isa<CXXConstructorDecl>(CGF.CurFuncDecl) ||
           isa<CXXDestructorDecl>(CGF.CurFuncDecl) ||
           (NewExceptionBehavior == llvm::fp::ebIgnore &&
            NewRoundingBehavior == llvm::RoundingMode::NearestTiesToEven)) &&
          "FPConstrained should be enabled on entire function");
 
-  auto mergeFnAttrValue = [&](StringRef Name, bool Value) {
+  auto mergeFnAttrValue = [&](llvm::AttributeKey Name, bool Value) {
     auto OldValue =
         CGF.CurFn->getFnAttribute(Name).getValueAsBool();
     auto NewValue = OldValue & Value;
     if (OldValue != NewValue)
       CGF.CurFn->addFnAttr(Name, llvm::toStringRef(NewValue));
   };
-  mergeFnAttrValue("no-infs-fp-math", FPFeatures.getNoHonorInfs());
-  mergeFnAttrValue("no-nans-fp-math", FPFeatures.getNoHonorNaNs());
-  mergeFnAttrValue("no-signed-zeros-fp-math", FPFeatures.getNoSignedZero());
-  mergeFnAttrValue("unsafe-fp-math", FPFeatures.getAllowFPReassociate() &&
-                                         FPFeatures.getAllowReciprocal() &&
-                                         FPFeatures.getAllowApproxFunc() &&
-                                         FPFeatures.getNoSignedZero());
+  mergeFnAttrValue(llvm::NoInfsFPMathAttr, FPFeatures.getNoHonorInfs());
+  mergeFnAttrValue(llvm::NoNansFPMathAttr, FPFeatures.getNoHonorNaNs());
+  mergeFnAttrValue(llvm::NoSignedZerosFPMathAttr, FPFeatures.getNoSignedZero());
+  mergeFnAttrValue(
+      llvm::UnsafeFPMathAttr,
+      FPFeatures.getAllowFPReassociate() && FPFeatures.getAllowReciprocal() &&
+          FPFeatures.getAllowApproxFunc() && FPFeatures.getNoSignedZero());
 }
 
 CodeGenFunction::CGFPOptionsRAII::~CGFPOptionsRAII() {
   CGF.CurFPFeatures = OldFPFeatures;
   CGF.Builder.setDefaultConstrainedExcept(OldExcept);
   CGF.Builder.setDefaultConstrainedRounding(OldRounding);
 }
 
if (HasCleanups) {
 
     PopCleanupBlocks(PrologueCleanupDepth);
   }
 
   // Emit function epilog (to return).
   llvm::DebugLoc Loc = EmitReturnBlock();
 
   if (ShouldInstrumentFunction()) {
-    if (CGM.getCodeGenOpts().InstrumentFunctions)
-      CurFn->addFnAttr("instrument-function-exit", "__cyg_profile_func_exit");
-    if (CGM.getCodeGenOpts().InstrumentFunctionsAfterInlining)
-      CurFn->addFnAttr("instrument-function-exit-inlined",
+    if (CGM.getCodeGenOpts().InstrumentFunctions) {
+      CurFn->addFnAttr(llvm::InstrumentFunctionExitAttr,
                        "__cyg_profile_func_exit");
     }
+    if (CGM.getCodeGenOpts().InstrumentFunctionsAfterInlining) {
+      CurFn->addFnAttr(llvm::InstrumentFunctionExitInlinedAttr,
+                       "__cyg_profile_func_exit");
+    }
+  }
 
   if (ShouldSkipSanitizerInstrumentation())
     CurFn->addFnAttr(llvm::Attribute::DisableSanitizerInstrumentation);
 
   // Emit debug descriptor for function end.
   if (CGDebugInfo *DI = getDebugInfo())
     DI->EmitFunctionEnd(Builder, CurFn);
 
void CodeGenFunction::FinishFunction(SourceLocation EndLoc) {
 
   // Add the required-vector-width attribute. This contains the max width from:
   // 1. min-vector-width attribute used in the source program.
   // 2. Any builtins used that have a vector width specified.
   // 3. Values passed in and out of inline assembly.
   // 4. Width of vector arguments and return types for this function.
   // 5. Width of vector aguments and return types for functions called by this
   //    function.
-  CurFn->addFnAttr("min-legal-vector-width", llvm::utostr(LargestVectorWidth));
+  CurFn->addFnAttr(llvm::MinLegalVectorWidthAttr,
+                   llvm::utostr(LargestVectorWidth));
 
   // Add vscale_range attribute if appropriate.
   Optional<std::pair<unsigned, unsigned>> VScaleRange =
       getContext().getTargetInfo().getVScaleRange(getLangOpts());
   if (VScaleRange) {
     CurFn->addFnAttr(llvm::Attribute::getWithVScaleRangeArgs(
         getLLVMContext(), VScaleRange.getValue().first,
         VScaleRange.getValue().second));
if (auto *CS = dyn_cast_or_null<CompoundStmt>(Body)) {
     if (LastStmt != CS->body_rend())
       return isa<ReturnStmt>(*LastStmt);
   }
   return false;
 }
 
 void CodeGenFunction::markAsIgnoreThreadCheckingAtRuntime(llvm::Function *Fn) {
   if (SanOpts.has(SanitizerKind::Thread)) {
-    Fn->addFnAttr("sanitize_thread_no_checking_at_run_time");
+    Fn->addFnAttr(llvm::SanitizeThreadNoCheckingAtRunTimeAttr);
     Fn->removeFnAttr(llvm::Attribute::SanitizeThread);
   }
 }
 
 /// Check if the return value of this function requires sanitization.
 bool CodeGenFunction::requiresReturnValueCheck() const {
   return requiresReturnValueNullabilityCheck() ||
          (SanOpts.has(SanitizerKind::ReturnsNonnullAttribute) && CurCodeDecl &&
#undef SANITIZER
   // Apply xray attributes to the function (as a string, for now)
   bool AlwaysXRayAttr = false;
   if (const auto *XRayAttr = D ? D->getAttr<XRayInstrumentAttr>() : nullptr) {
     if (CGM.getCodeGenOpts().XRayInstrumentationBundle.has(
             XRayInstrKind::FunctionEntry) ||
         CGM.getCodeGenOpts().XRayInstrumentationBundle.has(
             XRayInstrKind::FunctionExit)) {
       if (XRayAttr->alwaysXRayInstrument() && ShouldXRayInstrumentFunction()) {
-        Fn->addFnAttr("function-instrument", "xray-always");
+        Fn->addFnAttr(llvm::FunctionInstrumentAttr, "xray-always");
         AlwaysXRayAttr = true;
       }
       if (XRayAttr->neverXRayInstrument())
-        Fn->addFnAttr("function-instrument", "xray-never");
+        Fn->addFnAttr(llvm::FunctionInstrumentAttr, "xray-never");
       if (const auto *LogArgs = D->getAttr<XRayLogArgsAttr>())
-        if (ShouldXRayInstrumentFunction())
-          Fn->addFnAttr("xray-log-args",
+        if (ShouldXRayInstrumentFunction()) {
+          Fn->addFnAttr(llvm::XrayLogArgsAttr,
                         llvm::utostr(LogArgs->getArgumentCount()));
         }
+    }
   } else {
-    if (ShouldXRayInstrumentFunction() && !CGM.imbueXRayAttrs(Fn, Loc))
+    if (ShouldXRayInstrumentFunction() && !CGM.imbueXRayAttrs(Fn, Loc)) {
       Fn->addFnAttr(
-          "xray-instruction-threshold",
+          llvm::XrayInstructionThresholdAttr,
           llvm::itostr(CGM.getCodeGenOpts().XRayInstructionThreshold));
     }
+  }
 
   if (ShouldXRayInstrumentFunction()) {
-    if (CGM.getCodeGenOpts().XRayIgnoreLoops)
-      Fn->addFnAttr("xray-ignore-loops");
+    if (CGM.getCodeGenOpts().XRayIgnoreLoops) {
+      Fn->addFnAttr(llvm::XrayIgnoreLoopsAttr);
+    }
 
     if (!CGM.getCodeGenOpts().XRayInstrumentationBundle.has(
-            XRayInstrKind::FunctionExit))
-      Fn->addFnAttr("xray-skip-exit");
+            XRayInstrKind::FunctionExit)) {
+      Fn->addFnAttr(llvm::XraySkipExitAttr);
+    }
 
     if (!CGM.getCodeGenOpts().XRayInstrumentationBundle.has(
-            XRayInstrKind::FunctionEntry))
-      Fn->addFnAttr("xray-skip-entry");
+            XRayInstrKind::FunctionEntry)) {
+      Fn->addFnAttr(llvm::XraySkipEntryAttr);
+    }
 
     auto FuncGroups = CGM.getCodeGenOpts().XRayTotalFunctionGroups;
     if (FuncGroups > 1) {
       auto FuncName = llvm::makeArrayRef<uint8_t>(
           CurFn->getName().bytes_begin(), CurFn->getName().bytes_end());
       auto Group = crc32(FuncName) % FuncGroups;
       if (Group != CGM.getCodeGenOpts().XRaySelectedFunctionGroup &&
-          !AlwaysXRayAttr)
-        Fn->addFnAttr("function-instrument", "xray-never");
+          !AlwaysXRayAttr) {
+        Fn->addFnAttr(llvm::FunctionInstrumentAttr, "xray-never");
+      }
     }
   }
 
   if (CGM.getCodeGenOpts().getProfileInstr() != CodeGenOptions::ProfileNone)
     if (CGM.isProfileInstrExcluded(Fn, Loc))
       Fn->addFnAttr(llvm::Attribute::NoProfile);
 
   unsigned Count, Offset;
   if (const auto *Attr =
           D ? D->getAttr<PatchableFunctionEntryAttr>() : nullptr) {
     Count = Attr->getCount();
     Offset = Attr->getOffset();
   } else {
     Count = CGM.getCodeGenOpts().PatchableFunctionEntryCount;
     Offset = CGM.getCodeGenOpts().PatchableFunctionEntryOffset;
   }
   if (Count && Offset <= Count) {
-    Fn->addFnAttr("patchable-function-entry", std::to_string(Count - Offset));
-    if (Offset)
-      Fn->addFnAttr("patchable-function-prefix", std::to_string(Offset));
+    Fn->addFnAttr(llvm::PatchableFunctionEntryAttr,
+                  std::to_string(Count - Offset));
+    if (Offset) {
+      Fn->addFnAttr(llvm::PatchableFunctionPrefixAttr, std::to_string(Offset));
+    }
   }
 
   // Add no-jump-tables value.
-  if (CGM.getCodeGenOpts().NoUseJumpTables)
-    Fn->addFnAttr("no-jump-tables", "true");
+  if (CGM.getCodeGenOpts().NoUseJumpTables) {
+    Fn->addFnAttr(llvm::NoJumpTablesAttr, "true");
+  }
 
   // Add no-inline-line-tables value.
-  if (CGM.getCodeGenOpts().NoInlineLineTables)
-    Fn->addFnAttr("no-inline-line-tables");
+  if (CGM.getCodeGenOpts().NoInlineLineTables) {
+    Fn->addFnAttr(llvm::NoInlineLineTablesAttr);
+  }
 
   // Add profile-sample-accurate value.
-  if (CGM.getCodeGenOpts().ProfileSampleAccurate)
-    Fn->addFnAttr("profile-sample-accurate");
+  if (CGM.getCodeGenOpts().ProfileSampleAccurate) {
+    Fn->addFnAttr(llvm::ProfileSampleAccurateAttr);
+  }
 
-  if (!CGM.getCodeGenOpts().SampleProfileFile.empty())
-    Fn->addFnAttr("use-sample-profile");
+  if (!CGM.getCodeGenOpts().SampleProfileFile.empty()) {
+    Fn->addFnAttr(llvm::UseSampleProfileAttr);
+  }
 
-  if (D && D->hasAttr<CFICanonicalJumpTableAttr>())
-    Fn->addFnAttr("cfi-canonical-jump-table");
+  if (D && D->hasAttr<CFICanonicalJumpTableAttr>()) {
+    Fn->addFnAttr(llvm::CfiCanonicalJumpTablesAttr);
+  }
 
   if (D && D->hasAttr<NoProfileFunctionAttr>())
     Fn->addFnAttr(llvm::Attribute::NoProfile);
 
   if (FD && getLangOpts().OpenCL) {
     // Add metadata for a kernel function.
     EmitOpenCLKernelMetadata(FD, Fn);
   }
if ((FD && (FD->UsesFPIntrin() || FD->hasAttr<StrictFPAttr>())) ||
                RM != llvm::RoundingMode::NearestTiesToEven))) {
     Builder.setIsFPConstrained(true);
     Fn->addFnAttr(llvm::Attribute::StrictFP);
   }
 
   // If a custom alignment is used, force realigning to this alignment on
   // any main function which certainly will need it.
   if (FD && ((FD->isMain() || FD->isMSVCRTEntryPoint()) &&
-             CGM.getCodeGenOpts().StackAlignment))
-    Fn->addFnAttr("stackrealign");
+             CGM.getCodeGenOpts().StackAlignment)) {
+    Fn->addFnAttr(llvm::StackrealignAttr);
+  }
 
   llvm::BasicBlock *EntryBB = createBasicBlock("entry", CurFn);
 
   // Create a marker to make it easy to insert allocas into the entryblock
   // later.  Don't create this with the builder, because we don't want it
   // folded.
   llvm::Value *Undef = llvm::UndefValue::get(Int32Ty);
   AllocaInsertPt = new llvm::BitCastInst(Undef, Int32Ty, "allocapt", EntryBB);
if (CGDebugInfo *DI = getDebugInfo()) {
     // such as 'this' and 'vtt', show up in the debug info. Preserve the calling
     // convention.
     DI->emitFunctionStart(GD, Loc, StartLoc,
                           DI->getFunctionType(FD, RetTy, Args), CurFn,
                           CurFuncIsThunk);
   }
 
   if (ShouldInstrumentFunction()) {
-    if (CGM.getCodeGenOpts().InstrumentFunctions)
-      CurFn->addFnAttr("instrument-function-entry", "__cyg_profile_func_enter");
-    if (CGM.getCodeGenOpts().InstrumentFunctionsAfterInlining)
-      CurFn->addFnAttr("instrument-function-entry-inlined",
+    if (CGM.getCodeGenOpts().InstrumentFunctions) {
+      CurFn->addFnAttr(llvm::InstrumentFunctionEntryAttr,
+                       "__cyg_profile_func_enter");
+    }
+    if (CGM.getCodeGenOpts().InstrumentFunctionsAfterInlining) {
+      CurFn->addFnAttr(llvm::InstrumentFunctionEntryInlinedAttr,
                        "__cyg_profile_func_enter");
-    if (CGM.getCodeGenOpts().InstrumentFunctionEntryBare)
-      CurFn->addFnAttr("instrument-function-entry-inlined",
+    }
+    if (CGM.getCodeGenOpts().InstrumentFunctionEntryBare) {
+      CurFn->addFnAttr(llvm::InstrumentFunctionEntryInlinedAttr,
                        "__cyg_profile_func_enter_bare");
     }
+  }
 
   // Since emitting the mcount call here impacts optimizations such as function
   // inlining, we just add an attribute to insert a mcount call in backend.
   // The attribute "counting-function" is set to mcount function name which is
   // architecture dependent.
   if (CGM.getCodeGenOpts().InstrumentForProfiling) {
     // Calls to fentry/mcount should not be generated if function has
     // the no_instrument_function attribute.
     if (!CurFuncDecl || !CurFuncDecl->hasAttr<NoInstrumentFunctionAttr>()) {
-      if (CGM.getCodeGenOpts().CallFEntry)
-        Fn->addFnAttr("fentry-call", "true");
-      else {
-        Fn->addFnAttr("instrument-function-entry-inlined",
+      if (CGM.getCodeGenOpts().CallFEntry) {
+        Fn->addFnAttr(llvm::FentryCallAttr, "true");
+      } else {
+        Fn->addFnAttr(llvm::InstrumentFunctionEntryInlinedAttr,
                       getTarget().getMCountName());
       }
+
       if (CGM.getCodeGenOpts().MNopMCount) {
         if (!CGM.getCodeGenOpts().CallFEntry)
           CGM.getDiags().Report(diag::err_opt_not_valid_without_opt)
             << "-mnop-mcount" << "-mfentry";
-        Fn->addFnAttr("mnop-mcount");
+        Fn->addFnAttr(llvm::MnopMcountAttr);
       }
 
       if (CGM.getCodeGenOpts().RecordMCount) {
         if (!CGM.getCodeGenOpts().CallFEntry)
           CGM.getDiags().Report(diag::err_opt_not_valid_without_opt)
             << "-mrecord-mcount" << "-mfentry";
-        Fn->addFnAttr("mrecord-mcount");
+        Fn->addFnAttr(llvm::MrecordMcountAttr);
       }
     }
   }
 
   if (CGM.getCodeGenOpts().PackedStack) {
     if (getContext().getTargetInfo().getTriple().getArch() !=
         llvm::Triple::systemz)
       CGM.getDiags().Report(diag::err_opt_not_valid_on_target)
         << "-mpacked-stack";
-    Fn->addFnAttr("packed-stack");
+    Fn->addFnAttr(llvm::PackedStackAttr);
   }
 
   if (CGM.getCodeGenOpts().WarnStackSize != UINT_MAX &&
-      !CGM.getDiags().isIgnored(diag::warn_fe_backend_frame_larger_than, Loc))
-    Fn->addFnAttr("warn-stack-size",
+      !CGM.getDiags().isIgnored(diag::warn_fe_backend_frame_larger_than, Loc)) {
+    Fn->addFnAttr(llvm::WarnStackSizeAttr,
                   std::to_string(CGM.getCodeGenOpts().WarnStackSize));
+  }
 
   if (RetTy->isVoidType()) {
     // Void type; nothing to return.
     ReturnValue = Address::invalid();
 
     // Count the implicit return.
     if (!endsWithReturn(D))
       ++NumReturnExprs;

clang/lib/CodeGen/CodeGenModule.cpp

 
 void CodeGenModule::SetLLVMFunctionAttributesForDefinition(const Decl *D,
                                                            llvm::Function *F) {
   llvm::AttrBuilder B;
 
   if (CodeGenOpts.UnwindTables)
     B.addAttribute(llvm::Attribute::UWTable);
 
-  if (CodeGenOpts.StackClashProtector)
-    B.addAttribute("probe-stack", "inline-asm");
+  if (CodeGenOpts.StackClashProtector) {
+    B.addAttribute(llvm::ProbeStackAttr, "inline-asm");
+  }
 
   if (!hasUnwindExceptions(LangOpts))
     B.addAttribute(llvm::Attribute::NoUnwind);
 
   if (!D || !D->hasAttr<NoStackProtectorAttr>()) {
     if (LangOpts.getStackProtector() == LangOptions::SSPOn)
       B.addAttribute(llvm::Attribute::StackProtect);
     else if (LangOpts.getStackProtector() == LangOptions::SSPStrong)

       F->addTypeMetadata(0, Id);
     }
   }
 }
 
 void CodeGenModule::setLLVMFunctionFEnvAttributes(const FunctionDecl *D,
                                                   llvm::Function *F) {
   if (D->hasAttr<StrictFPAttr>()) {
-    llvm::AttrBuilder FuncAttrs;
-    FuncAttrs.addAttribute("strictfp");
-    F->addFnAttrs(FuncAttrs);
+    F->addFnAttr(llvm::StrictfpAttr);
   }
 }
 
 void CodeGenModule::SetCommonAttributes(GlobalDecl GD, llvm::GlobalValue *GV) {
   const Decl *D = GD.getDecl();
   if (dyn_cast_or_null<NamedDecl>(D))
     setGVProperties(GV, GD);
   else

     }
   } else {
     // Otherwise just add the existing target cpu and target features to the
     // function.
     Features = getTarget().getTargetOpts().Features;
   }
 
   if (!TargetCPU.empty()) {
-    Attrs.addAttribute("target-cpu", TargetCPU);
+    Attrs.addAttribute(llvm::TargetCPUAttr, TargetCPU);
     AddedAttr = true;
   }
   if (!TuneCPU.empty()) {
-    Attrs.addAttribute("tune-cpu", TuneCPU);
+    Attrs.addAttribute(llvm::TuneCPUAttr, TuneCPU);
     AddedAttr = true;
   }
   if (!Features.empty()) {
     llvm::sort(Features);
-    Attrs.addAttribute("target-features", llvm::join(Features, ","));
+    Attrs.addAttribute(llvm::TargetFeaturesAttr, llvm::join(Features, ","));
     AddedAttr = true;
   }
 
   return AddedAttr;
 }
 
 void CodeGenModule::setNonAliasAttributes(GlobalDecl GD,
                                           llvm::GlobalObject *GO) {
   const Decl *D = GD.getDecl();
   SetCommonAttributes(GD, GO);
 
   if (D) {
     if (auto *GV = dyn_cast<llvm::GlobalVariable>(GO)) {
       if (D->hasAttr<RetainAttr>())
         addUsedGlobal(GV);
-      if (auto *SA = D->getAttr<PragmaClangBSSSectionAttr>())
-        GV->addAttribute("bss-section", SA->getName());
-      if (auto *SA = D->getAttr<PragmaClangDataSectionAttr>())
-        GV->addAttribute("data-section", SA->getName());
-      if (auto *SA = D->getAttr<PragmaClangRodataSectionAttr>())
-        GV->addAttribute("rodata-section", SA->getName());
-      if (auto *SA = D->getAttr<PragmaClangRelroSectionAttr>())
-        GV->addAttribute("relro-section", SA->getName());
+      if (auto *SA = D->getAttr<PragmaClangBSSSectionAttr>()) {
+        GV->addAttribute(llvm::BssSectionAttr, SA->getName());
+      }
+      if (auto *SA = D->getAttr<PragmaClangDataSectionAttr>()) {
+        GV->addAttribute(llvm::DataSectionAttr, SA->getName());
+      }
+      if (auto *SA = D->getAttr<PragmaClangRodataSectionAttr>()) {
+        GV->addAttribute(llvm::RodataSectionAttr, SA->getName());
+      }
+      if (auto *SA = D->getAttr<PragmaClangRelroSectionAttr>()) {
+        GV->addAttribute(llvm::RelroSectionAttr, SA->getName());
+      }
     }
 
     if (auto *F = dyn_cast<llvm::Function>(GO)) {
       if (D->hasAttr<RetainAttr>())
         addUsedGlobal(F);
       if (auto *SA = D->getAttr<PragmaClangTextSectionAttr>())
-        if (!D->getAttr<SectionAttr>())
-          F->addFnAttr("implicit-section-name", SA->getName());
+        if (!D->getAttr<SectionAttr>()) {
+          F->addFnAttr(llvm::ImplicitSectionNameAttr, SA->getName());
+        }
 
       llvm::AttrBuilder Attrs;
       if (GetCPUAndFeaturesAttributes(GD, Attrs)) {
         // We know that GetCPUAndFeaturesAttributes will always have the
         // newest set, since it has the newest possible FunctionDecl, so the
         // new ones should replace the old.
         llvm::AttrBuilder RemoveAttrs;
-        RemoveAttrs.addAttribute("target-cpu");
-        RemoveAttrs.addAttribute("target-features");
-        RemoveAttrs.addAttribute("tune-cpu");
+        RemoveAttrs.addAttribute(llvm::TargetCPUAttr);
+        RemoveAttrs.addAttribute(llvm::TargetFeaturesAttr);
+        RemoveAttrs.addAttribute(llvm::TuneCPUAttr);
         F->removeFnAttrs(RemoveAttrs);
         F->addFnAttrs(Attrs);
       }
     }
 
     if (const auto *CSA = D->getAttr<CodeSegAttr>())
       GO->setSection(CSA->getName());
     else if (const auto *SA = D->getAttr<SectionAttr>())

     getTargetCodeGenInfo().setTargetAttributes(FD, F, *this);
 
   if (const auto *CSA = FD->getAttr<CodeSegAttr>())
     F->setSection(CSA->getName());
   else if (const auto *SA = FD->getAttr<SectionAttr>())
      F->setSection(SA->getName());
 
   if (const auto *EA = FD->getAttr<ErrorAttr>()) {
-    if (EA->isError())
-      F->addFnAttr("dontcall-error", EA->getUserDiagnostic());
-    else if (EA->isWarning())
-      F->addFnAttr("dontcall-warn", EA->getUserDiagnostic());
+    if (EA->isError()) {
+      F->addFnAttr(llvm::DontcallErrorAttr, EA->getUserDiagnostic());
+    } else if (EA->isWarning()) {
+      F->addFnAttr(llvm::DontcallWarnAttr, EA->getUserDiagnostic());
+    }
   }
 
   // If we plan on emitting this inline builtin, we can't treat it as a builtin.
   if (FD->isInlineBuiltinDeclaration()) {
     const FunctionDecl *FDBody;
     bool HasBody = FD->hasBody(FDBody);
     (void)HasBody;
     assert(HasBody && "Inline builtin declarations should always have an "

   if (Loc.isValid())
     Attr = XRayFilter.shouldImbueLocation(Loc, Category);
   if (Attr == ImbueAttr::NONE)
     Attr = XRayFilter.shouldImbueFunction(Fn->getName());
   switch (Attr) {
   case ImbueAttr::NONE:
     return false;
   case ImbueAttr::ALWAYS:
-    Fn->addFnAttr("function-instrument", "xray-always");
+    Fn->addFnAttr(llvm::FunctionInstrumentAttr, "xray-always");
     break;
   case ImbueAttr::ALWAYS_ARG1:
-    Fn->addFnAttr("function-instrument", "xray-always");
-    Fn->addFnAttr("xray-log-args", "1");
+    Fn->addFnAttr(llvm::FunctionInstrumentAttr, "xray-always");
+    Fn->addFnAttr(llvm::XrayLogArgsAttr, "1");
     break;
   case ImbueAttr::NEVER:
-    Fn->addFnAttr("function-instrument", "xray-never");
+    Fn->addFnAttr(llvm::FunctionInstrumentAttr, "xray-never");
     break;
   }
   return true;
 }
 
 bool CodeGenModule::isProfileInstrExcluded(llvm::Function *Fn,
                                            SourceLocation Loc) const {
   const auto &ProfileList = getContext().getProfileList();

   // properly aligned on 32-bit platforms.
   CharUnits Alignment =
       IsSwiftABI ? Context.toCharUnitsFromBits(64) : getPointerAlign();
 
   // The struct.
   GV = Fields.finishAndCreateGlobal("_unnamed_cfstring_", Alignment,
                                     /*isConstant=*/false,
                                     llvm::GlobalVariable::PrivateLinkage);
-  GV->addAttribute("objc_arc_inert");
+  GV->addAttribute(llvm::ObjcArcInertAttr);
   switch (Triple.getObjectFormat()) {
   case llvm::Triple::UnknownObjectFormat:
     llvm_unreachable("unknown file format");
   case llvm::Triple::GOFF:
     llvm_unreachable("GOFF is not yet implemented");
   case llvm::Triple::XCOFF:
     llvm_unreachable("XCOFF is not yet implemented");
   case llvm::Triple::COFF:

clang/lib/CodeGen/MicrosoftCXXABI.cpp

 
   CGM.SetLLVMFunctionAttributes(MD, FnInfo, ThunkFn, /*IsThunk=*/false);
   CGM.SetLLVMFunctionAttributesForDefinition(MD, ThunkFn);
 
   // Add the "thunk" attribute so that LLVM knows that the return type is
   // meaningless. These thunks can be used to call functions with differing
   // return types, and the caller is required to cast the prototype
   // appropriately to extract the correct value.
-  ThunkFn->addFnAttr("thunk");
+  ThunkFn->addFnAttr(llvm::ThunkAttr);
 
   // These thunks can be compared, so they are not unnamed.
   ThunkFn->setUnnamedAddr(llvm::GlobalValue::UnnamedAddr::None);
 
   // Start codegen.
   CodeGenFunction CGF(CGM);
   CGF.CurGD = GlobalDecl(MD);
   CGF.CurFuncIsThunk = true;

clang/lib/CodeGen/TargetInfo.cpp

 
   void setTargetAttributes(const Decl *D, llvm::GlobalValue *GV,
                            CodeGen::CodeGenModule &CGM) const override {
     TargetCodeGenInfo::setTargetAttributes(D, GV, CGM);
     if (const auto *FD = dyn_cast_or_null<FunctionDecl>(D)) {
       if (const auto *Attr = FD->getAttr<WebAssemblyImportModuleAttr>()) {
         llvm::Function *Fn = cast<llvm::Function>(GV);
         llvm::AttrBuilder B;
-        B.addAttribute("wasm-import-module", Attr->getImportModule());
+        B.addAttribute(llvm::WasmImportModuleAttr, Attr->getImportModule());
         Fn->addFnAttrs(B);
       }
       if (const auto *Attr = FD->getAttr<WebAssemblyImportNameAttr>()) {
         llvm::Function *Fn = cast<llvm::Function>(GV);
         llvm::AttrBuilder B;
-        B.addAttribute("wasm-import-name", Attr->getImportName());
+        B.addAttribute(llvm::WasmImportNameAttr, Attr->getImportName());
         Fn->addFnAttrs(B);
       }
       if (const auto *Attr = FD->getAttr<WebAssemblyExportNameAttr>()) {
         llvm::Function *Fn = cast<llvm::Function>(GV);
         llvm::AttrBuilder B;
-        B.addAttribute("wasm-export-name", Attr->getExportName());
+        B.addAttribute(llvm::WasmExportNameAttr, Attr->getExportName());
         Fn->addFnAttrs(B);
       }
     }
 
     if (auto *FD = dyn_cast_or_null<FunctionDecl>(D)) {
       llvm::Function *Fn = cast<llvm::Function>(GV);
-      if (!FD->doesThisDeclarationHaveABody() && !FD->hasPrototype())
-        Fn->addFnAttr("no-prototype");
+      if (!FD->doesThisDeclarationHaveABody() && !FD->hasPrototype()) {
+        Fn->addFnAttr(llvm::NoPrototypeAttr);
+      }
     }
   }
 };
 
 /// Classify argument of given type \p Ty.
 ABIArgInfo WebAssemblyABIInfo::classifyArgumentType(QualType Ty) const {
   Ty = useFirstFieldIfTransparentUnion(Ty);
 

 
 void X86_32TargetCodeGenInfo::setTargetAttributes(
     const Decl *D, llvm::GlobalValue *GV, CodeGen::CodeGenModule &CGM) const {
   if (GV->isDeclaration())
     return;
   if (const FunctionDecl *FD = dyn_cast_or_null<FunctionDecl>(D)) {
     if (FD->hasAttr<X86ForceAlignArgPointerAttr>()) {
       llvm::Function *Fn = cast<llvm::Function>(GV);
-      Fn->addFnAttr("stackrealign");
+      Fn->addFnAttr(llvm::StackrealignAttr);
     }
 
     addX86InterruptAttrs(FD, GV, CGM);
   }
 }
 
 bool X86_32TargetCodeGenInfo::initDwarfEHRegSizeTable(
                                                CodeGen::CodeGenFunction &CGF,

 
   void setTargetAttributes(const Decl *D, llvm::GlobalValue *GV,
                            CodeGen::CodeGenModule &CGM) const override {
     if (GV->isDeclaration())
       return;
     if (const FunctionDecl *FD = dyn_cast_or_null<FunctionDecl>(D)) {
       if (FD->hasAttr<X86ForceAlignArgPointerAttr>()) {
         llvm::Function *Fn = cast<llvm::Function>(GV);
-        Fn->addFnAttr("stackrealign");
+        Fn->addFnAttr(llvm::StackrealignAttr);
       }
 
       addX86InterruptAttrs(FD, GV, CGM);
     }
   }
 
   void checkFunctionCallABI(CodeGenModule &CGM, SourceLocation CallLoc,
                             const FunctionDecl *Caller,

     Opt = "/FAILIFMISMATCH:\"" + Name.str() + "=" + Value.str() + "\"";
   }
 };
 
 static void addStackProbeTargetAttributes(const Decl *D, llvm::GlobalValue *GV,
                                           CodeGen::CodeGenModule &CGM) {
   if (llvm::Function *Fn = dyn_cast_or_null<llvm::Function>(GV)) {
 
-    if (CGM.getCodeGenOpts().StackProbeSize != 4096)
-      Fn->addFnAttr("stack-probe-size",
+    if (CGM.getCodeGenOpts().StackProbeSize != 4096) {
+      Fn->addFnAttr(llvm::StackProbeSizeAttr,
                     llvm::utostr(CGM.getCodeGenOpts().StackProbeSize));
-    if (CGM.getCodeGenOpts().NoStackArgProbe)
-      Fn->addFnAttr("no-stack-arg-probe");
+    }
+    if (CGM.getCodeGenOpts().NoStackArgProbe) {
+      Fn->addFnAttr(llvm::NoStackArgProbeAttr);
+    }
   }
 }
 
 void WinX86_32TargetCodeGenInfo::setTargetAttributes(
     const Decl *D, llvm::GlobalValue *GV, CodeGen::CodeGenModule &CGM) const {
   X86_32TargetCodeGenInfo::setTargetAttributes(D, GV, CGM);
   if (GV->isDeclaration())
     return;

 void WinX86_64TargetCodeGenInfo::setTargetAttributes(
     const Decl *D, llvm::GlobalValue *GV, CodeGen::CodeGenModule &CGM) const {
   TargetCodeGenInfo::setTargetAttributes(D, GV, CGM);
   if (GV->isDeclaration())
     return;
   if (const FunctionDecl *FD = dyn_cast_or_null<FunctionDecl>(D)) {
     if (FD->hasAttr<X86ForceAlignArgPointerAttr>()) {
       llvm::Function *Fn = cast<llvm::Function>(GV);
-      Fn->addFnAttr("stackrealign");
+      Fn->addFnAttr(llvm::StackrealignAttr);
     }
 
     addX86InterruptAttrs(FD, GV, CGM);
   }
 
   addStackProbeTargetAttributes(D, GV, CGM);
 }
 }

     TargetInfo::BranchProtectionInfo BPI;
     StringRef Error;
     (void)CGM.getTarget().validateBranchProtection(Attr.BranchProtection,
                                                    BPI, Error);
     assert(Error.empty());
 
     auto *Fn = cast<llvm::Function>(GV);
     static const char *SignReturnAddrStr[] = {"none", "non-leaf", "all"};
-    Fn->addFnAttr("sign-return-address", SignReturnAddrStr[static_cast<int>(BPI.SignReturnAddr)]);
+    Fn->addFnAttr(llvm::SignReturnAddressAttr,
+                  SignReturnAddrStr[static_cast<int>(BPI.SignReturnAddr)]);
 
     if (BPI.SignReturnAddr != LangOptions::SignReturnAddressScopeKind::None) {
-      Fn->addFnAttr("sign-return-address-key",
+      Fn->addFnAttr(llvm::SignReturnAddressAttr,
                     BPI.SignKey == LangOptions::SignReturnAddressKeyKind::AKey
                         ? "a_key"
                         : "b_key");
     }
 
-    Fn->addFnAttr("branch-target-enforcement",
+    Fn->addFnAttr(llvm::BranchTargetEnforcementAttr,
                   BPI.BranchTargetEnforcement ? "true" : "false");
   }
 
   bool isScalarizableAsmOperand(CodeGen::CodeGenFunction &CGF,
                                 llvm::Type *Ty) const override {
     if (CGF.getTarget().hasFeature("ls64")) {
       auto *ST = dyn_cast<llvm::StructType>(Ty);
       if (ST && ST->getNumElements() == 1) {

     case ARMInterruptAttr::FIQ:     Kind = "FIQ"; break;
     case ARMInterruptAttr::SWI:     Kind = "SWI"; break;
     case ARMInterruptAttr::ABORT:   Kind = "ABORT"; break;
     case ARMInterruptAttr::UNDEF:   Kind = "UNDEF"; break;
     }
 
     llvm::Function *Fn = cast<llvm::Function>(GV);
 
-    Fn->addFnAttr("interrupt", Kind);
+    Fn->addFnAttr(llvm::InterruptAttr, Kind);
 
     ARMABIInfo::ABIKind ABI = cast<ARMABIInfo>(getABIInfo()).getABIKind();
     if (ABI == ARMABIInfo::APCS)
       return;
 
     // AAPCS guarantees that sp will be 8-byte aligned on any public interface,
     // however this is not necessarily true on taking any interrupt. Instruct
     // the backend to perform a realignment as part of the function prologue.

 
 }
 
 void MSP430TargetCodeGenInfo::setTargetAttributes(
     const Decl *D, llvm::GlobalValue *GV, CodeGen::CodeGenModule &M) const {
   if (GV->isDeclaration())
     return;
   if (const FunctionDecl *FD = dyn_cast_or_null<FunctionDecl>(D)) {
-    const auto *InterruptAttr = FD->getAttr<MSP430InterruptAttr>();
-    if (!InterruptAttr)
+    const auto *Interrupt_Attr = FD->getAttr<MSP430InterruptAttr>();
+    if (!Interrupt_Attr)
       return;
 
     // Handle 'interrupt' attribute:
     llvm::Function *F = cast<llvm::Function>(GV);
 
     // Step 1: Set ISR calling convention.
     F->setCallingConv(llvm::CallingConv::MSP430_INTR);
 
     // Step 2: Add attributes goodness.
     F->addFnAttr(llvm::Attribute::NoInline);
-    F->addFnAttr("interrupt", llvm::utostr(InterruptAttr->getNumber()));
+    F->addFnAttr(llvm::InterruptAttr,
+                 llvm::utostr(Interrupt_Attr->getNumber()));
   }
 }
 
 //===----------------------------------------------------------------------===//
 // MIPS ABI Implementation.  This works for both little-endian and
 // big-endian variants.
 //===----------------------------------------------------------------------===//
 

   }
 
   void setTargetAttributes(const Decl *D, llvm::GlobalValue *GV,
                            CodeGen::CodeGenModule &CGM) const override {
     const FunctionDecl *FD = dyn_cast_or_null<FunctionDecl>(D);
     if (!FD) return;
     llvm::Function *Fn = cast<llvm::Function>(GV);
 
-    if (FD->hasAttr<MipsLongCallAttr>())
-      Fn->addFnAttr("long-call");
-    else if (FD->hasAttr<MipsShortCallAttr>())
-      Fn->addFnAttr("short-call");
+    if (FD->hasAttr<MipsLongCallAttr>()) {
+      Fn->addFnAttr(llvm::LongCallAttr);
+    } else if (FD->hasAttr<MipsShortCallAttr>()) {
+      Fn->addFnAttr(llvm::ShortCallAttr);
+    }
 
     // Other attributes do not have a meaning for declarations.
     if (GV->isDeclaration())
       return;
 
     if (FD->hasAttr<Mips16Attr>()) {
-      Fn->addFnAttr("mips16");
+      Fn->addFnAttr(llvm::Mips16KAttr);
     }
     else if (FD->hasAttr<NoMips16Attr>()) {
-      Fn->addFnAttr("nomips16");
+      Fn->addFnAttr(llvm::Nomips16Attr);
     }
 
-    if (FD->hasAttr<MicroMipsAttr>())
-      Fn->addFnAttr("micromips");
-    else if (FD->hasAttr<NoMicroMipsAttr>())
-      Fn->addFnAttr("nomicromips");
+    if (FD->hasAttr<MicroMipsAttr>()) {
+      Fn->addFnAttr(llvm::MicromipsAttr);
+    } else if (FD->hasAttr<NoMicroMipsAttr>()) {
+      Fn->addFnAttr(llvm::NoMicromipsAttr);
+    }
 
     const MipsInterruptAttr *Attr = FD->getAttr<MipsInterruptAttr>();
     if (!Attr)
       return;
 
     const char *Kind;
     switch (Attr->getInterrupt()) {
     case MipsInterruptAttr::eic:     Kind = "eic"; break;
     case MipsInterruptAttr::sw0:     Kind = "sw0"; break;
     case MipsInterruptAttr::sw1:     Kind = "sw1"; break;
     case MipsInterruptAttr::hw0:     Kind = "hw0"; break;
     case MipsInterruptAttr::hw1:     Kind = "hw1"; break;
     case MipsInterruptAttr::hw2:     Kind = "hw2"; break;
     case MipsInterruptAttr::hw3:     Kind = "hw3"; break;
     case MipsInterruptAttr::hw4:     Kind = "hw4"; break;
     case MipsInterruptAttr::hw5:     Kind = "hw5"; break;
     }
 
-    Fn->addFnAttr("interrupt", Kind);
-
+    Fn->addFnAttr(llvm::InterruptAttr, Kind);
   }
 
   bool initDwarfEHRegSizeTable(CodeGen::CodeGenFunction &CGF,
                                llvm::Value *Address) const override;
 
   unsigned getSizeOfUnwindException() const override {
     return SizeOfUnwindException;
   }

   void setTargetAttributes(const Decl *D, llvm::GlobalValue *GV,
                            CodeGen::CodeGenModule &CGM) const override {
     if (GV->isDeclaration())
       return;
     const auto *FD = dyn_cast_or_null<FunctionDecl>(D);
     if (!FD) return;
     auto *Fn = cast<llvm::Function>(GV);
 
-    if (FD->getAttr<AVRInterruptAttr>())
-      Fn->addFnAttr("interrupt");
+    if (FD->getAttr<AVRInterruptAttr>()) {
+      Fn->addFnAttr(llvm::InterruptAttr);
+    }
 
-    if (FD->getAttr<AVRSignalAttr>())
-      Fn->addFnAttr("signal");
+    if (FD->getAttr<AVRSignalAttr>()) {
+      Fn->addFnAttr(llvm::SignalAttr);
+    }
   }
 };
 }
 
 //===----------------------------------------------------------------------===//
 // TCE ABI Implementation (see http://tce.cs.tut.fi). Uses mostly the defaults.
 // Currently subclassed only to implement custom OpenCL C function attribute
 // handling.

     FD->getAttr<ReqdWorkGroupSizeAttr>() : nullptr;
 
 
   const bool IsOpenCLKernel = M.getLangOpts().OpenCL &&
                               FD->hasAttr<OpenCLKernelAttr>();
   const bool IsHIPKernel = M.getLangOpts().HIP &&
                            FD->hasAttr<CUDAGlobalAttr>();
   if ((IsOpenCLKernel || IsHIPKernel) &&
-      (M.getTriple().getOS() == llvm::Triple::AMDHSA))
-    F->addFnAttr("amdgpu-implicitarg-num-bytes", "56");
-
-  if (IsHIPKernel)
-    F->addFnAttr("uniform-work-group-size", "true");
+      (M.getTriple().getOS() == llvm::Triple::AMDHSA)) {
+    F->addFnAttr(llvm::AmdgpuImplicitargNumBytesAttr, "56");
+  }
 
+  if (IsHIPKernel) {
+    F->addFnAttr(llvm::UniformWorkGroupSizeAttr, "true");
+  }
 
   const auto *FlatWGS = FD->getAttr<AMDGPUFlatWorkGroupSizeAttr>();
   if (ReqdWGS || FlatWGS) {
     unsigned Min = 0;
     unsigned Max = 0;
     if (FlatWGS) {
       Min = FlatWGS->getMin()
                 ->EvaluateKnownConstInt(M.getContext())
                 .getExtValue();
       Max = FlatWGS->getMax()
                 ->EvaluateKnownConstInt(M.getContext())
                 .getExtValue();
     }
     if (ReqdWGS && Min == 0 && Max == 0)
       Min = Max = ReqdWGS->getXDim() * ReqdWGS->getYDim() * ReqdWGS->getZDim();
 
     if (Min != 0) {
       assert(Min <= Max && "Min must be less than or equal Max");
 
       std::string AttrVal = llvm::utostr(Min) + "," + llvm::utostr(Max);
-      F->addFnAttr("amdgpu-flat-work-group-size", AttrVal);
+      F->addFnAttr(llvm::AmdgpuFlatWorkGroupSizeAttr, AttrVal);
     } else
       assert(Max == 0 && "Max must be zero");
   } else if (IsOpenCLKernel || IsHIPKernel) {
     // By default, restrict the maximum size to a value specified by
     // --gpu-max-threads-per-block=n or its default value for HIP.
     const unsigned OpenCLDefaultMaxWorkGroupSize = 256;
     const unsigned DefaultMaxWorkGroupSize =
         IsOpenCLKernel ? OpenCLDefaultMaxWorkGroupSize
                        : M.getLangOpts().GPUMaxThreadsPerBlock;
     std::string AttrVal =
         std::string("1,") + llvm::utostr(DefaultMaxWorkGroupSize);
-    F->addFnAttr("amdgpu-flat-work-group-size", AttrVal);
+    F->addFnAttr(llvm::AmdgpuFlatWorkGroupSizeAttr, AttrVal);
   }
 
   if (const auto *Attr = FD->getAttr<AMDGPUWavesPerEUAttr>()) {
     unsigned Min =
         Attr->getMin()->EvaluateKnownConstInt(M.getContext()).getExtValue();
     unsigned Max = Attr->getMax() ? Attr->getMax()
                                         ->EvaluateKnownConstInt(M.getContext())
                                         .getExtValue()
                                   : 0;
 
     if (Min != 0) {
       assert((Max == 0 || Min <= Max) && "Min must be less than or equal Max");
 
       std::string AttrVal = llvm::utostr(Min);
       if (Max != 0)
         AttrVal = AttrVal + "," + llvm::utostr(Max);
-      F->addFnAttr("amdgpu-waves-per-eu", AttrVal);
+      F->addFnAttr(llvm::AmdgpuWavesPerEuAttr, AttrVal);
     } else
       assert(Max == 0 && "Max must be zero");
   }
 
   if (const auto *Attr = FD->getAttr<AMDGPUNumSGPRAttr>()) {
     unsigned NumSGPR = Attr->getNumSGPR();
 
-    if (NumSGPR != 0)
-      F->addFnAttr("amdgpu-num-sgpr", llvm::utostr(NumSGPR));
+    if (NumSGPR != 0) {
+      F->addFnAttr(llvm::AmdgpuNumSgprAttr, llvm::utostr(NumSGPR));
+    }
   }
 
   if (const auto *Attr = FD->getAttr<AMDGPUNumVGPRAttr>()) {
     uint32_t NumVGPR = Attr->getNumVGPR();
 
-    if (NumVGPR != 0)
-      F->addFnAttr("amdgpu-num-vgpr", llvm::utostr(NumVGPR));
+    if (NumVGPR != 0) {
+      F->addFnAttr(llvm::AmdgpuNumVgprAttr, llvm::utostr(NumVGPR));
+    }
   }
 
-  if (M.getContext().getTargetInfo().allowAMDGPUUnsafeFPAtomics())
-    F->addFnAttr("amdgpu-unsafe-fp-atomics", "true");
+  if (M.getContext().getTargetInfo().allowAMDGPUUnsafeFPAtomics()) {
+    F->addFnAttr(llvm::AmdgpuUnsafeFpAtomicsAttr, "true");
+  }
 
-  if (!getABIInfo().getCodeGenOpts().EmitIEEENaNCompliantInsts)
-    F->addFnAttr("amdgpu-ieee", "false");
+  if (!getABIInfo().getCodeGenOpts().EmitIEEENaNCompliantInsts) {
+    F->addFnAttr(llvm::AmdgpuIeeeAttr, "false");
+  }
 }
 
 unsigned AMDGPUTargetCodeGenInfo::getOpenCLKernelCallingConv() const {
   return llvm::CallingConv::AMDGPU_KERNEL;
 }
 
 // Currently LLVM assumes null pointers always have value 0,
 // which results in incorrectly transformed IR. Therefore, instead of

     switch (Attr->getInterrupt()) {
     case RISCVInterruptAttr::user: Kind = "user"; break;
     case RISCVInterruptAttr::supervisor: Kind = "supervisor"; break;
     case RISCVInterruptAttr::machine: Kind = "machine"; break;
     }
 
     auto *Fn = cast<llvm::Function>(GV);
 
-    Fn->addFnAttr("interrupt", Kind);
+    Fn->addFnAttr(llvm::InterruptAttr, Kind);
   }
 };
 } // namespace
 
 //===----------------------------------------------------------------------===//
 // VE ABI Implementation.
 //
 namespace {

     ArgTypeQuals.push_back(llvm::MDString::get(C, ""));
     ArgNames.push_back(
         llvm::MDString::get(C, (Twine("local_arg") + Twine(I)).str()));
   }
   std::string Name = Invoke->getName().str() + "_kernel";
   auto *FT = llvm::FunctionType::get(llvm::Type::getVoidTy(C), ArgTys, false);
   auto *F = llvm::Function::Create(FT, llvm::GlobalValue::InternalLinkage, Name,
                                    &CGF.CGM.getModule());
-  F->addFnAttr("enqueued-block");
+  F->addFnAttr(llvm::EnqueuedBlockAttr);
   auto IP = CGF.Builder.saveIP();
   auto *BB = llvm::BasicBlock::Create(C, "entry", F);
   Builder.SetInsertPoint(BB);
   const auto BlockAlign = CGF.CGM.getDataLayout().getPrefTypeAlign(BlockTy);
   auto *BlockPtr = Builder.CreateAlloca(BlockTy, nullptr);
   BlockPtr->setAlignment(BlockAlign);
   Builder.CreateAlignedStore(F->arg_begin(), BlockPtr, BlockAlign);
   auto *Cast = Builder.CreatePointerCast(BlockPtr, InvokeFT->getParamType(0));

llvm/include/llvm/Analysis/SyncDependenceAnalysis.h

 
 #ifndef LLVM_ANALYSIS_SYNCDEPENDENCEANALYSIS_H
 #define LLVM_ANALYSIS_SYNCDEPENDENCEANALYSIS_H
 
 #include "llvm/ADT/DenseMap.h"
 #include "llvm/ADT/PostOrderIterator.h"
 #include "llvm/ADT/SmallPtrSet.h"
 #include "llvm/Analysis/LoopInfo.h"
+#include <map>
 #include <memory>
 #include <unordered_map>
 
 namespace llvm {
 
 class BasicBlock;
 class DominatorTree;
 class Loop;

llvm/include/llvm/Analysis/TargetLibraryInfo.h

class TargetLibraryInfo {
   BitVector OverrideAsUnavailable;
 
 public:
   explicit TargetLibraryInfo(const TargetLibraryInfoImpl &Impl,
                              Optional<const Function *> F = None)
       : Impl(&Impl), OverrideAsUnavailable(NumLibFuncs) {
     if (!F)
       return;
-    if ((*F)->hasFnAttribute("no-builtins"))
+    if ((*F)->hasFnAttribute(NoBuiltinsAttr))
       disableAllFunctions();
     else {
       // Disable individual libc/libm calls in TargetLibraryInfo.
       LibFunc LF;
       AttributeSet FnAttrs = (*F)->getAttributes().getFnAttrs();
       for (const Attribute &Attr : FnAttrs) {
         if (!Attr.isStringAttribute())
           continue;
-        auto AttrStr = Attr.getKindAsString();
+        auto AttrStr = Attr.getKindAsKey().value();
         if (!AttrStr.consume_front("no-builtin-"))
           continue;
         if (getLibFunc(AttrStr, LF))
           setUnavailable(LF);
       }
     }
   }
 

llvm/include/llvm/Analysis/TargetTransformInfoImpl.h

void getMemcpyLoopResidualLoweringType(
       unsigned RemainingBytes, unsigned SrcAddrSpace, unsigned DestAddrSpace,
       unsigned SrcAlign, unsigned DestAlign) const {
     for (unsigned i = 0; i != RemainingBytes; ++i)
       OpsOut.push_back(Type::getInt8Ty(Context));
   }
 
   bool areInlineCompatible(const Function *Caller,
                            const Function *Callee) const {
-    return (Caller->getFnAttribute("target-cpu") ==
-            Callee->getFnAttribute("target-cpu")) &&
-           (Caller->getFnAttribute("target-features") ==
-            Callee->getFnAttribute("target-features"));
+    return (Caller->getFnAttribute(TargetCPUAttr) ==
+            Callee->getFnAttribute(TargetCPUAttr)) &&
+           (Caller->getFnAttribute(TargetFeaturesAttr) ==
+            Callee->getFnAttribute(TargetFeaturesAttr));
   }
 
   bool areFunctionArgsABICompatible(const Function *Caller,
                                     const Function *Callee,
                                     SmallPtrSetImpl<Argument *> &Args) const {
-    return (Caller->getFnAttribute("target-cpu") ==
-            Callee->getFnAttribute("target-cpu")) &&
-           (Caller->getFnAttribute("target-features") ==
-            Callee->getFnAttribute("target-features"));
+    return (Caller->getFnAttribute(TargetCPUAttr) ==
+            Callee->getFnAttribute(TargetCPUAttr)) &&
+           (Caller->getFnAttribute(TargetFeaturesAttr) ==
+            Callee->getFnAttribute(TargetFeaturesAttr));
   }
 
   bool isIndexedLoadLegal(TTI::MemIndexedMode Mode, Type *Ty,
                           const DataLayout &DL) const {
     return false;
   }
 
   bool isIndexedStoreLegal(TTI::MemIndexedMode Mode, Type *Ty,

llvm/include/llvm/Analysis/VectorUtils.h

 /// <scalarname> = the name of the scalar function.
 /// <vectorname> = the name of the vector function.
 std::string mangleTLIVectorName(StringRef VectorName, StringRef ScalarName,
                                 unsigned numArgs, ElementCount VF);
 
 /// Retrieve the `VFParamKind` from a string token.
 VFParamKind getVFParamKindFromString(const StringRef Token);
 
-// Name of the attribute where the variant mappings are stored.
-static constexpr char const *MappingsAttrName = "vector-function-abi-variant";
-
 /// Populates a set of strings representing the Vector Function ABI variants
 /// associated to the CallInst CI. If the CI does not contain the
 /// vector-function-abi-variant attribute, we return without populating
 /// VariantMappings, i.e. callers of getVectorVariantNames need not check for
 /// the presence of the attribute (see InjectTLIMappings).
 void getVectorVariantNames(const CallInst &CI,
                            SmallVectorImpl<std::string> &VariantMappings);
 } // end namespace VFABI

llvm/include/llvm/CodeGen/CommandFlags.h

 TargetOptions InitTargetOptionsFromCodeGenFlags(const llvm::Triple &TheTriple);
 
 std::string getCPUStr();
 
 std::string getFeaturesStr();
 
 std::vector<std::string> getFeatureList();
 
-void renderBoolStringAttr(AttrBuilder &B, StringRef Name, bool Val);
+void renderBoolStringAttr(AttrBuilder &B, AttributeKey Name, bool Val);
 
 /// Set function attributes of function \p F based on CPU, Features, and command
 /// line flags.
 void setFunctionAttributes(StringRef CPU, StringRef Features, Function &F);
 
 /// Set function attributes of functions in Module M based on CPU,
 /// Features, and command line flags.
 void setFunctionAttributes(StringRef CPU, StringRef Features, Module &M);
 } // namespace codegen
 } // namespace llvm
 
 #endif // LLVM_CODEGEN_COMMANDFLAGS_H

llvm/include/llvm/CodeGen/TargetLowering.h

   /// Return true if the operation uses custom lowering, regardless of whether
   /// the type is legal or not.
   bool isOperationCustom(unsigned Op, EVT VT) const {
     return getOperationAction(Op, VT) == Custom;
   }
 
   /// Return true if lowering to a jump table is allowed.
   virtual bool areJTsAllowed(const Function *Fn) const {
-    if (Fn->getFnAttribute("no-jump-tables").getValueAsBool())
+    if (Fn->getFnAttribute(NoJumpTablesAttr).getValueAsBool())
       return false;
 
     return isOperationLegalOrCustom(ISD::BR_JT, MVT::Other) ||
            isOperationLegalOrCustom(ISD::BRIND, MVT::Other);
   }
 
   /// Check whether the range [Low,High] fits in a machine word.
   bool rangeFitsInWord(const APInt &Low, const APInt &High,

llvm/include/llvm/IR/Assumptions.h

 //===----------------------------------------------------------------------===//
 
 #ifndef LLVM_IR_ASSUMPTIONS_H
 #define LLVM_IR_ASSUMPTIONS_H
 
 #include "llvm/ADT/DenseSet.h"
 #include "llvm/ADT/StringRef.h"
 #include "llvm/ADT/StringSet.h"
+#include "llvm/IR/Attributes.h"
 
 namespace llvm {
 
 class Function;
 class CallBase;
 
-/// The key we use for assumption attributes.
-constexpr StringRef AssumptionAttrKey = "llvm.assume";
-
 /// A set of known assumption strings that are accepted without warning and
 /// which can be recommended as typo correction.
 extern StringSet<> KnownAssumptionStrings;
 
 /// Helper that allows to insert a new assumption string in the known assumption
 /// set by creating a (static) object.
 struct KnownAssumptionString {
   KnownAssumptionString(StringRef AssumptionStr)

llvm/include/llvm/IR/Attributes.h

 //
 //===----------------------------------------------------------------------===//
 
 #ifndef LLVM_IR_ATTRIBUTES_H
 #define LLVM_IR_ATTRIBUTES_H
 
 #include "llvm-c/Types.h"
 #include "llvm/ADT/ArrayRef.h"
+#include "llvm/ADT/DenseMap.h"
 #include "llvm/ADT/Optional.h"
 #include "llvm/ADT/SmallString.h"
 #include "llvm/ADT/StringRef.h"
 #include "llvm/ADT/iterator_range.h"
 #include "llvm/Config/llvm-config.h"
 #include "llvm/Support/Alignment.h"
 #include "llvm/Support/PointerLikeTypeTraits.h"
 #include <bitset>
 #include <cassert>
 #include <cstdint>
 #include <map>
 #include <string>
+#include <unordered_map>
 #include <utility>
+namespace llvm {
+constexpr uint32_t dj2(const char *str) {
+  uint32_t hash = 5381;
+  int c = 0;
+
+  while ((c = *str++))
+    hash = ((hash << 5) + hash) + c; /* hash * 33 + c */
+
+  return hash;
+}
+
+class AttributeKey {
+  const char *value_;
+  uint32_t hash_;
+  uint32_t size_;
+
+  AttributeKey(StringRef s)
+      : value_(strndup(s.data(), s.size())), hash_(dj2(s.data())),
+        size_(s.size()) {}
+
+public:
+  static AttributeKey Create(StringRef s);
+
+  template <std::size_t N>
+  explicit constexpr AttributeKey(const char (&s)[N])
+      : value_(s), hash_(dj2(s)), size_(N - 1) {
+    static_assert(N - 1 < std::numeric_limits<decltype(size_)>::max(),
+                  "large enough size storage");
+  }
+  constexpr AttributeKey() : AttributeKey("") {}
+  StringRef value() const { return StringRef(value_, size_); }
+  uint32_t hash() const { return hash_; }
+  uint32_t size() const { return size_; }
+  bool operator==(AttributeKey const &other) const {
+    return size_ == other.size_ && strncmp(value_, other.value_, size_) == 0;
+  }
+  bool operator!=(AttributeKey const &other) const {
+    return !((*this) == other);
+  }
+  bool operator<(AttributeKey const &other) const {
+    return strcmp(value_, other.value_) < 0;
+  }
+};
+} // namespace llvm
+
+template <> struct std::hash<llvm::AttributeKey> {
+  std::size_t operator()(llvm::AttributeKey const &s) const noexcept {
+    return s.hash();
+  }
+};
 
 namespace llvm {
 
 class AttrBuilder;
 class AttributeImpl;
 class AttributeListImpl;
 class AttributeSetNode;
 template<typename T> struct DenseMapInfo;
public:
   Attribute() = default;
 
   //===--------------------------------------------------------------------===//
   // Attribute Construction
   //===--------------------------------------------------------------------===//
 
   /// Return a uniquified Attribute object.
   static Attribute get(LLVMContext &Context, AttrKind Kind, uint64_t Val = 0);
-  static Attribute get(LLVMContext &Context, StringRef Kind,
+  static Attribute get(LLVMContext &Context, AttributeKey Kind,
                        StringRef Val = StringRef());
   static Attribute get(LLVMContext &Context, AttrKind Kind, Type *Ty);
 
   /// Return a uniquified Attribute object that has the specific
   /// alignment set.
   static Attribute getWithAlignment(LLVMContext &Context, Align Alignment);
   static Attribute getWithStackAlignment(LLVMContext &Context, Align Alignment);
   static Attribute getWithDereferenceableBytes(LLVMContext &Context,
public:
 
   /// Return true if the attribute is any kind of attribute.
   bool isValid() const { return pImpl; }
 
   /// Return true if the attribute is present.
   bool hasAttribute(AttrKind Val) const;
 
   /// Return true if the target-dependent attribute is present.
-  bool hasAttribute(StringRef Val) const;
+  bool hasAttribute(AttributeKey Val) const;
 
   /// Return the attribute's kind as an enum (Attribute::AttrKind). This
   /// requires the attribute to be an enum, integer, or type attribute.
   Attribute::AttrKind getKindAsEnum() const;
 
   /// Return the attribute's value as an integer. This requires that the
   /// attribute be an integer attribute.
   uint64_t getValueAsInt() const;
 
   /// Return the attribute's value as a boolean. This requires that the
   /// attribute be a string attribute.
   bool getValueAsBool() const;
 
   /// Return the attribute's kind as a string. This requires the
   /// attribute to be a string attribute.
-  StringRef getKindAsString() const;
+  AttributeKey getKindAsKey() const;
 
   /// Return the attribute's value as a string. This requires the
   /// attribute to be a string attribute.
   StringRef getValueAsString() const;
 
   /// Return the attribute's value as a Type. This requires the attribute to be
   /// a type attribute.
   Type *getValueAsType() const;
public:
 
   /// Add an argument attribute. Returns a new set because attribute sets are
   /// immutable.
   LLVM_NODISCARD AttributeSet addAttribute(LLVMContext &C,
                                            Attribute::AttrKind Kind) const;
 
   /// Add a target-dependent attribute. Returns a new set because attribute sets
   /// are immutable.
-  LLVM_NODISCARD AttributeSet addAttribute(LLVMContext &C, StringRef Kind,
+  LLVM_NODISCARD AttributeSet addAttribute(LLVMContext &C, AttributeKey Kind,
                                            StringRef Value = StringRef()) const;
 
   /// Add attributes to the attribute set. Returns a new set because attribute
   /// sets are immutable.
   LLVM_NODISCARD AttributeSet addAttributes(LLVMContext &C,
                                             AttributeSet AS) const;
 
   /// Remove the specified attribute from this set. Returns a new set because
   /// attribute sets are immutable.
   LLVM_NODISCARD AttributeSet removeAttribute(LLVMContext &C,
                                               Attribute::AttrKind Kind) const;
 
   /// Remove the specified attribute from this set. Returns a new set because
   /// attribute sets are immutable.
   LLVM_NODISCARD AttributeSet removeAttribute(LLVMContext &C,
-                                              StringRef Kind) const;
+                                              AttributeKey Kind) const;
 
   /// Remove the specified attributes from this set. Returns a new set because
   /// attribute sets are immutable.
   LLVM_NODISCARD AttributeSet
   removeAttributes(LLVMContext &C, const AttrBuilder &AttrsToRemove) const;
 
   /// Return the number of attributes in this set.
   unsigned getNumAttributes() const;
 
   /// Return true if attributes exists in this set.
   bool hasAttributes() const { return SetNode != nullptr; }
 
   /// Return true if the attribute exists in this set.
   bool hasAttribute(Attribute::AttrKind Kind) const;
 
   /// Return true if the attribute exists in this set.
-  bool hasAttribute(StringRef Kind) const;
+  bool hasAttribute(AttributeKey Kind) const;
 
   /// Return the attribute object.
   Attribute getAttribute(Attribute::AttrKind Kind) const;
 
   /// Return the target-dependent attribute object.
-  Attribute getAttribute(StringRef Kind) const;
+  Attribute getAttribute(AttributeKey Kind) const;
 
   MaybeAlign getAlignment() const;
   MaybeAlign getStackAlignment() const;
   uint64_t getDereferenceableBytes() const;
   uint64_t getDereferenceableOrNullBytes() const;
   Type *getByValType() const;
   Type *getStructRetType() const;
   Type *getByRefType() const;
public:
   /// Return an AttributeList with the specified parameters in it.
   static AttributeList get(LLVMContext &C, ArrayRef<AttributeList> Attrs);
   static AttributeList get(LLVMContext &C, unsigned Index,
                            ArrayRef<Attribute::AttrKind> Kinds);
   static AttributeList get(LLVMContext &C, unsigned Index,
                            ArrayRef<Attribute::AttrKind> Kinds,
                            ArrayRef<uint64_t> Values);
   static AttributeList get(LLVMContext &C, unsigned Index,
-                           ArrayRef<StringRef> Kind);
+                           ArrayRef<AttributeKey> Kind);
   static AttributeList get(LLVMContext &C, unsigned Index,
                            const AttrBuilder &B);
 
   // TODO: remove non-AtIndex versions of these methods.
   /// Add an attribute to the attribute set at the given index.
   /// Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList addAttributeAtIndex(
       LLVMContext &C, unsigned Index, Attribute::AttrKind Kind) const;
 
   /// Add an attribute to the attribute set at the given index.
   /// Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList
-  addAttributeAtIndex(LLVMContext &C, unsigned Index, StringRef Kind,
+  addAttributeAtIndex(LLVMContext &C, unsigned Index, AttributeKey Kind,
                       StringRef Value = StringRef()) const;
 
   /// Add an attribute to the attribute set at the given index.
   /// Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList addAttributeAtIndex(LLVMContext &C,
                                                    unsigned Index,
                                                    Attribute A) const;
 
public:
   LLVM_NODISCARD AttributeList addFnAttribute(LLVMContext &C,
                                               Attribute Attr) const {
     return addAttributeAtIndex(C, FunctionIndex, Attr);
   }
 
   /// Add a function attribute to the list. Returns a new list because
   /// attribute lists are immutable.
   LLVM_NODISCARD AttributeList addFnAttribute(
-      LLVMContext &C, StringRef Kind, StringRef Value = StringRef()) const {
+      LLVMContext &C, AttributeKey Kind, StringRef Value = StringRef()) const {
     return addAttributeAtIndex(C, FunctionIndex, Kind, Value);
   }
 
   /// Add function attribute to the list. Returns a new list because
   /// attribute lists are immutable.
   LLVM_NODISCARD AttributeList addFnAttributes(LLVMContext &C,
                                                const AttrBuilder &B) const {
     return addAttributesAtIndex(C, FunctionIndex, B);
public:
   LLVM_NODISCARD AttributeList addParamAttribute(
       LLVMContext &C, unsigned ArgNo, Attribute::AttrKind Kind) const {
     return addAttributeAtIndex(C, ArgNo + FirstArgIndex, Kind);
   }
 
   /// Add an argument attribute to the list. Returns a new list because
   /// attribute lists are immutable.
   LLVM_NODISCARD AttributeList
-  addParamAttribute(LLVMContext &C, unsigned ArgNo, StringRef Kind,
+  addParamAttribute(LLVMContext &C, unsigned ArgNo, AttributeKey Kind,
                     StringRef Value = StringRef()) const {
     return addAttributeAtIndex(C, ArgNo + FirstArgIndex, Kind, Value);
   }
 
   /// Add an attribute to the attribute list at the given arg indices. Returns a
   /// new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList addParamAttribute(LLVMContext &C,
                                                  ArrayRef<unsigned> ArgNos,
public:
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeAttributeAtIndex(
       LLVMContext &C, unsigned Index, Attribute::AttrKind Kind) const;
 
   /// Remove the specified attribute at the specified index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeAttributeAtIndex(LLVMContext &C,
                                                       unsigned Index,
-                                                      StringRef Kind) const;
+                                                      AttributeKey Kind) const;
   LLVM_NODISCARD AttributeList removeAttribute(LLVMContext &C, unsigned Index,
-                                               StringRef Kind) const {
+                                               AttributeKey Kind) const {
     return removeAttributeAtIndex(C, Index, Kind);
   }
 
   /// Remove the specified attributes at the specified index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeAttributesAtIndex(
       LLVMContext &C, unsigned Index, const AttrBuilder &AttrsToRemove) const;
 
   /// Remove all attributes at the specified index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeAttributesAtIndex(LLVMContext &C,
                                                        unsigned Index) const;
 
   /// Remove the specified attribute at the function index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList
   removeFnAttribute(LLVMContext &C, Attribute::AttrKind Kind) const {
     return removeAttributeAtIndex(C, FunctionIndex, Kind);
   }
 
   /// Remove the specified attribute at the function index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeFnAttribute(LLVMContext &C,
-                                                 StringRef Kind) const {
+                                                 AttributeKey Kind) const {
     return removeAttributeAtIndex(C, FunctionIndex, Kind);
   }
 
   /// Remove the specified attribute at the function index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList
   removeFnAttributes(LLVMContext &C, const AttrBuilder &AttrsToRemove) const {
     return removeAttributesAtIndex(C, FunctionIndex, AttrsToRemove);
public:
   LLVM_NODISCARD AttributeList
   removeRetAttribute(LLVMContext &C, Attribute::AttrKind Kind) const {
     return removeAttributeAtIndex(C, ReturnIndex, Kind);
   }
 
   /// Remove the specified attribute at the return value index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeRetAttribute(LLVMContext &C,
-                                                  StringRef Kind) const {
+                                                  AttributeKey Kind) const {
     return removeAttributeAtIndex(C, ReturnIndex, Kind);
   }
 
   /// Remove the specified attribute at the return value index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList
   removeRetAttributes(LLVMContext &C, const AttrBuilder &AttrsToRemove) const {
     return removeAttributesAtIndex(C, ReturnIndex, AttrsToRemove);
   }
 
   /// Remove the specified attribute at the specified arg index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeParamAttribute(
       LLVMContext &C, unsigned ArgNo, Attribute::AttrKind Kind) const {
     return removeAttributeAtIndex(C, ArgNo + FirstArgIndex, Kind);
   }
 
   /// Remove the specified attribute at the specified arg index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeParamAttribute(LLVMContext &C,
                                                     unsigned ArgNo,
-                                                    StringRef Kind) const {
+                                                    AttributeKey Kind) const {
     return removeAttributeAtIndex(C, ArgNo + FirstArgIndex, Kind);
   }
 
   /// Remove the specified attribute at the specified arg index from this
   /// attribute list. Returns a new list because attribute lists are immutable.
   LLVM_NODISCARD AttributeList removeParamAttributes(
       LLVMContext &C, unsigned ArgNo, const AttrBuilder &AttrsToRemove) const {
     return removeAttributesAtIndex(C, ArgNo + FirstArgIndex, AttrsToRemove);
public:
 
   /// The function attributes are returned.
   AttributeSet getFnAttrs() const;
 
   /// Return true if the attribute exists at the given index.
   bool hasAttributeAtIndex(unsigned Index, Attribute::AttrKind Kind) const;
 
   /// Return true if the attribute exists at the given index.
-  bool hasAttributeAtIndex(unsigned Index, StringRef Kind) const;
+  bool hasAttributeAtIndex(unsigned Index, AttributeKey Kind) const;
 
   /// Return true if attribute exists at the given index.
   bool hasAttributesAtIndex(unsigned Index) const;
 
   /// Return true if the attribute exists for the given argument
   bool hasParamAttr(unsigned ArgNo, Attribute::AttrKind Kind) const {
     return hasAttributeAtIndex(ArgNo + FirstArgIndex, Kind);
   }
 
   /// Return true if the attribute exists for the given argument
-  bool hasParamAttr(unsigned ArgNo, StringRef Kind) const {
+  bool hasParamAttr(unsigned ArgNo, AttributeKey Kind) const {
     return hasAttributeAtIndex(ArgNo + FirstArgIndex, Kind);
   }
 
   /// Return true if attributes exists for the given argument
   bool hasParamAttrs(unsigned ArgNo) const {
     return hasAttributesAtIndex(ArgNo + FirstArgIndex);
   }
 
   /// Return true if the attribute exists for the return value.
   bool hasRetAttr(Attribute::AttrKind Kind) const {
     return hasAttributeAtIndex(ReturnIndex, Kind);
   }
 
   /// Return true if the attribute exists for the return value.
-  bool hasRetAttr(StringRef Kind) const {
+  bool hasRetAttr(AttributeKey Kind) const {
     return hasAttributeAtIndex(ReturnIndex, Kind);
   }
 
   /// Return true if attributes exist for the return value.
   bool hasRetAttrs() const { return hasAttributesAtIndex(ReturnIndex); }
 
   /// Return true if the attribute exists for the function.
   bool hasFnAttr(Attribute::AttrKind Kind) const;
 
   /// Return true if the attribute exists for the function.
-  bool hasFnAttr(StringRef Kind) const;
+  bool hasFnAttr(AttributeKey Kind) const;
 
   /// Return true the attributes exist for the function.
   bool hasFnAttrs() const { return hasAttributesAtIndex(FunctionIndex); }
 
   /// Return true if the specified attribute is set for at least one
   /// parameter or for the return value. If Index is not nullptr, the index
   /// of a parameter with the specified attribute is provided.
   bool hasAttrSomewhere(Attribute::AttrKind Kind,
                         unsigned *Index = nullptr) const;
 
   /// Return the attribute object that exists at the given index.
   Attribute getAttributeAtIndex(unsigned Index, Attribute::AttrKind Kind) const;
 
   /// Return the attribute object that exists at the given index.
-  Attribute getAttributeAtIndex(unsigned Index, StringRef Kind) const;
+  Attribute getAttributeAtIndex(unsigned Index, AttributeKey Kind) const;
 
   /// Return the attribute object that exists at the arg index.
   Attribute getParamAttr(unsigned ArgNo, Attribute::AttrKind Kind) const {
     return getAttributeAtIndex(ArgNo + FirstArgIndex, Kind);
   }
 
   /// Return the attribute object that exists at the given index.
-  Attribute getParamAttr(unsigned ArgNo, StringRef Kind) const {
+  Attribute getParamAttr(unsigned ArgNo, AttributeKey Kind) const {
     return getAttributeAtIndex(ArgNo + FirstArgIndex, Kind);
   }
 
   /// Return the attribute object that exists for the function.
   Attribute getFnAttr(Attribute::AttrKind Kind) const {
     return getAttributeAtIndex(FunctionIndex, Kind);
   }
 
   /// Return the attribute object that exists for the function.
-  Attribute getFnAttr(StringRef Kind) const {
+  Attribute getFnAttr(AttributeKey Kind) const {
     return getAttributeAtIndex(FunctionIndex, Kind);
   }
 
   /// Return the alignment of the return value.
   MaybeAlign getRetAlignment() const;
 
   /// Return the alignment for the specified function parameter.
   MaybeAlign getParamAlignment(unsigned ArgNo) const;
return (unsigned((uintptr_t)AS.pImpl) >> 4) ^
            (unsigned((uintptr_t)AS.pImpl) >> 9);
   }
 
   static bool isEqual(AttributeList LHS, AttributeList RHS) {
     return LHS == RHS;
   }
 };
 
+constexpr llvm::AttributeKey
+    AmdgpuFlatWorkGroupSizeAttr("amdgpu-flat-work-group-size");
+constexpr llvm::AttributeKey AmdgpuIeeeAttr("amdgpu-ieee");
+constexpr llvm::AttributeKey
+    AmdgpuImplicitargNumBytesAttr("amdgpu-implicitarg-num-bytes");
+constexpr llvm::AttributeKey AmdgpuNumSgprAttr("amdgpu-num-sgpr");
+constexpr llvm::AttributeKey AmdgpuNumVgprAttr("amdgpu-num-vgpr");
+constexpr llvm::AttributeKey
+    AmdgpuUnsafeFpAtomicsAttr("amdgpu-unsafe-fp-atomics");
+constexpr llvm::AttributeKey AmdgpuWavesPerEuAttr("amdgpu-waves-per-eu");
+constexpr llvm::AttributeKey ApproxFuncFPMathAttr("approx-func-fp-math");
+constexpr llvm::AttributeKey BackchainAttr("backchain");
+constexpr llvm::AttributeKey
+    BranchTargetEnforcementAttr("branch-target-enforcement");
+constexpr llvm::AttributeKey BssSectionAttr("bss-section");
+constexpr llvm::AttributeKey CallThresholdBonus("call-threshold-bonus");
+constexpr llvm::AttributeKey CallInlineCost("call-inline-cost");
+constexpr llvm::AttributeKey
+    CfiCanonicalJumpTablesAttr("cfi-canonical-jump-table");
+constexpr llvm::AttributeKey CmseNonsecureEntryAttr("cmse_nonsecure_entry");
+constexpr llvm::AttributeKey CoroutinePresplitAttr("coroutine.presplit");
+constexpr llvm::AttributeKey DataSectionAttr("data-section");
+constexpr llvm::AttributeKey DenormalFPMathAttr("denormal-fp-math");
+constexpr llvm::AttributeKey DenormalFPMathf32Attr("denormal-fp-math-f32");
+constexpr llvm::AttributeKey DeoptLoweringAttr("deopt-lowering");
+constexpr llvm::AttributeKey DeviceInitAttr("device-init");
+constexpr llvm::AttributeKey DisableTailCallsAttr("disable-tail-calls");
+constexpr llvm::AttributeKey DontcallErrorAttr("dontcall-error");
+constexpr llvm::AttributeKey DontcallWarnAttr("dontcall-warn");
+constexpr llvm::AttributeKey EnqueuedBlockAttr("enqueued-block");
+constexpr llvm::AttributeKey FentryCallAttr("fentry-call");
+constexpr llvm::AttributeKey FramePointerAttr("frame-pointer");
+constexpr llvm::AttributeKey FunctionInlineCost("function-inline-cost");
+constexpr llvm::AttributeKey
+    FunctionInlineThreshold("function-inline-threshold");
+constexpr llvm::AttributeKey FunctionInstrumentAttr("function-instrument");
+constexpr llvm::AttributeKey GcLeafFunctionAttr("gc-leaf-function");
+constexpr llvm::AttributeKey GuardNocfAttr("guard_nocf");
+constexpr llvm::AttributeKey ImplicitSectionNameAttr("implicit-section-name");
+constexpr llvm::AttributeKey IndirectTLSSegRefsAttr("indirect-tls-seg-refs");
+constexpr llvm::AttributeKey InlineRemarkAttr("inline-remark");
+constexpr llvm::AttributeKey
+    InstrumentFunctionEntryAttr("instrument-function-entry");
+constexpr llvm::AttributeKey
+    InstrumentFunctionEntryInlinedAttr("instrument-function-entry-inlined");
+constexpr llvm::AttributeKey
+    InstrumentFunctionExitAttr("instrument-function-exit");
+constexpr llvm::AttributeKey
+    InstrumentFunctionExitInlinedAttr("instrument-function-exit-inlined");
+constexpr llvm::AttributeKey InterruptAttr("interrupt");
+constexpr llvm::AttributeKey LLVMAssumeAttr("llvm.assume");
+constexpr llvm::AttributeKey LessPreciseFPMADAttr("less-precise-fpmad");
+constexpr llvm::AttributeKey LongCallAttr("long-call");
+constexpr llvm::AttributeKey MicromipsAttr("micromips");
+constexpr llvm::AttributeKey MinLegalVectorWidthAttr("min-legal-vector-width");
+constexpr llvm::AttributeKey Mips16KAttr("mips16");
+constexpr llvm::AttributeKey MnopMcountAttr("mnop-mcount");
+constexpr llvm::AttributeKey MrecordMcountAttr("mrecord-mcount");
+constexpr llvm::AttributeKey MustProgressAttr("must-progress");
+constexpr llvm::AttributeKey NoBuiltinsAttr("no-builtins");
+constexpr llvm::AttributeKey
+    NoCalleeSavedRegistersAttr("no_callee_saved_registers");
+constexpr llvm::AttributeKey
+    NoCallerSavedRegistersAttr("no_caller_saved_registers");
+constexpr llvm::AttributeKey
+    NoCaptureMaybeReturned("no-capture-maybe-returned");
+constexpr llvm::AttributeKey NoFramePointerElimAttr("no-frame-pointer-elim");
+constexpr llvm::AttributeKey
+    NoFramePointerElimNonLeafAttr("no-frame-pointer-elim-non-leaf");
+constexpr llvm::AttributeKey NoImplicitFloatAttr("no-implicit-float");
+constexpr llvm::AttributeKey NoInfsFPMathAttr("no-infs-fp-math");
+constexpr llvm::AttributeKey NoInlineLineTablesAttr("no-inline-line-tables");
+constexpr llvm::AttributeKey NoJumpTablesAttr("no-jump-tables");
+constexpr llvm::AttributeKey NoOpenmp("no_openmp");
+constexpr llvm::AttributeKey NoOpenmpRoutines("no_openmp_routines");
+constexpr llvm::AttributeKey NoMicromipsAttr("no-micromips");
+constexpr llvm::AttributeKey NoNansFPMathAttr("no-nans-fp-math");
+constexpr llvm::AttributeKey NoProfileAttr("no-profile");
+constexpr llvm::AttributeKey NoPrototypeAttr("no-prototype");
+constexpr llvm::AttributeKey NoRealignStack("no-realign-stack");
+constexpr llvm::AttributeKey NoSignedZerosFPMathAttr("no-signed-zeros-fp-math");
+constexpr llvm::AttributeKey NoStackArgProbeAttr("no-stack-arg-probe");
+constexpr llvm::AttributeKey NoTrappingMathAttr("no-trapping-math");
+constexpr llvm::AttributeKey NomicromipsAttr("nomicromips");
+constexpr llvm::AttributeKey Nomips16Attr("nomips16");
+constexpr llvm::AttributeKey NullPointerIsValidAttr("null-pointer-is-valid");
+constexpr llvm::AttributeKey ObjcArcInertAttr("objc_arc_inert");
+constexpr llvm::AttributeKey OmpTargetNumTeamsAttr("omp_target_num_teams");
+constexpr llvm::AttributeKey
+    OmpTargetThreadLimitAttr("omp_target_thread_limit");
+constexpr llvm::AttributeKey PackedStackAttr("packed-stack");
+constexpr llvm::AttributeKey PatchableFunctionAttr("patchable-function");
+constexpr llvm::AttributeKey
+    PatchableFunctionEntryAttr("patchable-function-entry");
+constexpr llvm::AttributeKey
+    PatchableFunctionPrefixAttr("patchable-function-prefix");
+constexpr llvm::AttributeKey PreferVectorWidthAttr("prefer-vector-width");
+constexpr llvm::AttributeKey ProbeStackAttr("probe-stack");
+constexpr llvm::AttributeKey
+    ProfileSampleAccurateAttr("profile-sample-accurate");
+constexpr llvm::AttributeKey ReciprocalEstimatesAttr("reciprocal-estimates");
+constexpr llvm::AttributeKey RelroSectionAttr("relro-section");
+constexpr llvm::AttributeKey RodataSectionAttr("rodata-section");
+constexpr llvm::AttributeKey SafeStackAttr("safe-stack");
+constexpr llvm::AttributeKey SafesehAttr("safeseh");
+constexpr llvm::AttributeKey SampleProfileSuffixElisionPolicyAttr(
+    "sample-profile-suffix-elision-policy");
+constexpr llvm::AttributeKey SanitizeAddressAttr("sanitize-address");
+constexpr llvm::AttributeKey SanitizeHWAddressAttr("sanitize-hw-address");
+constexpr llvm::AttributeKey SanitizeMemTagAttr("sanitize-mem-tag");
+constexpr llvm::AttributeKey SanitizeMemoryAttr("sanitize-memory");
+constexpr llvm::AttributeKey SanitizeThreadAttr("sanitize-thread");
+constexpr llvm::AttributeKey SanitizeThreadNoCheckingAtRunTimeAttr(
+    "sanitize_thread_no_checking_at_run_time");
+constexpr llvm::AttributeKey ShadowCallStackAttr("shadow-call-stack");
+constexpr llvm::AttributeKey ShortCallAttr("short-call");
+constexpr llvm::AttributeKey SignReturnAddressAttr("sign-return-address");
+constexpr llvm::AttributeKey
+    SignReturnAddressKeyAttr("sign-return-address-key");
+constexpr llvm::AttributeKey SignalAttr("signal");
+constexpr llvm::AttributeKey
+    SpeculativeLoadHardeningAttr("speculative-load-hardening");
+constexpr llvm::AttributeKey SplitStackAttr("split-stack");
+constexpr llvm::AttributeKey StackProbeSizeAttr("stack-probe-size");
+constexpr llvm::AttributeKey
+    StackProtectorBufferSizeAttr("stack-protector-buffer-size");
+constexpr llvm::AttributeKey StackrealignAttr("stackrealign");
+constexpr llvm::AttributeKey StatepointIdAttr("statepoint-id");
+constexpr llvm::AttributeKey
+    StatepointNumPatchBytesAttr("statepoint-num-patch-bytes");
+constexpr llvm::AttributeKey
+    StrictFloatCastOverflowAttr("strict-float-cast-overflow");
+constexpr llvm::AttributeKey StrictfpAttr("strictfp");
+constexpr llvm::AttributeKey TargetCPUAttr("target-cpu");
+constexpr llvm::AttributeKey TargetFeaturesAttr("target-features");
+constexpr llvm::AttributeKey ThinltoInternalizeAttr("thinlto-internalize");
+constexpr llvm::AttributeKey ThunkAttr("thunk");
+constexpr llvm::AttributeKey TocDataAttr("toc-data");
+constexpr llvm::AttributeKey TrapFuncNameAttr("trap-func-name");
+constexpr llvm::AttributeKey TuneCPUAttr("tune-cpu");
+constexpr llvm::AttributeKey
+    UniformWorkGroupSizeAttr("uniform-work-group-size");
+constexpr llvm::AttributeKey UnsafeFPMathAttr("unsafe-fp-math");
+constexpr llvm::AttributeKey UseSampleProfileAttr("use-sample-profile");
+constexpr llvm::AttributeKey UseSoftFloatAttr("use-soft-float");
+constexpr llvm::AttributeKey
+    VectorFunctionAbiVariantAttr("vector-function-abi-variant");
+constexpr llvm::AttributeKey WarnStackSizeAttr("warn-stack-size");
+constexpr llvm::AttributeKey WasmExportNameAttr("wasm-export-name");
+constexpr llvm::AttributeKey WasmImportModuleAttr("wasm-import-module");
+constexpr llvm::AttributeKey WasmImportNameAttr("wasm-import-name");
+constexpr llvm::AttributeKey XrayIgnoreLoopsAttr("xray-ignore-loops");
+constexpr llvm::AttributeKey
+    XrayInstructionThresholdAttr("xray-instruction-threshold");
+constexpr llvm::AttributeKey XrayLogArgsAttr("xray-log-args");
+constexpr llvm::AttributeKey XraySkipEntryAttr("xray-skip-entry");
+constexpr llvm::AttributeKey XraySkipExitAttr("xray-skip-exit");
+
 //===----------------------------------------------------------------------===//
 /// \class
 /// This class is used in conjunction with the Attribute::get method to
 /// create an Attribute object. The object itself is uniquified. The Builder's
 /// value, however, is not. So this can be used as a quick way to test for
 /// equality, presence of attributes, etc.
 class AttrBuilder {
   std::bitset<Attribute::EndAttrKinds> Attrs;
-  std::map<SmallString<32>, SmallString<32>, std::less<>> TargetDepAttrs;
+  std::map<AttributeKey, SmallString<32>> TargetDepAttrs;
   std::array<uint64_t, Attribute::NumIntAttrKinds> IntAttrs = {};
   std::array<Type *, Attribute::NumTypeAttrKinds> TypeAttrs = {};
 
   Optional<unsigned> kindToIntIndex(Attribute::AttrKind Kind) const;
   Optional<unsigned> kindToTypeIndex(Attribute::AttrKind Kind) const;
 
 public:
   AttrBuilder() = default;
AttrBuilder &addAttribute(Attribute::AttrKind Val) {
     Attrs[Val] = true;
     return *this;
   }
 
   /// Add the Attribute object to the builder.
   AttrBuilder &addAttribute(Attribute A);
 
   /// Add the target-dependent attribute to the builder.
-  AttrBuilder &addAttribute(StringRef A, StringRef V = StringRef());
+  AttrBuilder &addAttribute(AttributeKey A, StringRef V = StringRef());
 
   /// Remove an attribute from the builder.
   AttrBuilder &removeAttribute(Attribute::AttrKind Val);
 
   /// Remove the attributes from the builder.
   AttrBuilder &removeAttributes(AttributeList A, uint64_t WithoutIndex);
 
   /// Remove the target-dependent attribute to the builder.
-  AttrBuilder &removeAttribute(StringRef A);
+  AttrBuilder &removeAttribute(AttributeKey A);
 
   /// Add the attributes from the builder.
   AttrBuilder &merge(const AttrBuilder &B);
 
   /// Remove the attributes from the builder.
   AttrBuilder &remove(const AttrBuilder &B);
 
   /// Return true if the builder has any attribute that's in the
   /// specified builder.
   bool overlaps(const AttrBuilder &B) const;
 
   /// Return true if the builder has the specified attribute.
   bool contains(Attribute::AttrKind A) const {
     assert((unsigned)A < Attribute::EndAttrKinds && "Attribute out of range!");
     return Attrs[A];
   }
 
   /// Return true if the builder has the specified target-dependent
   /// attribute.
-  bool contains(StringRef A) const;
+  bool contains(AttributeKey A) const;
 
   /// Return true if the builder has IR-level attributes.
   bool hasAttributes() const;
 
   /// Return true if the builder has any attribute that's in the
   /// specified attribute.
   bool hasAttributes(AttributeList A, uint64_t Index) const;
 

llvm/include/llvm/IR/Attributes.td

 /// Int attribute.
 class IntAttr<string S, list<AttrProperty> P> : Attr<S, P>;
 
 /// Type attribute.
 class TypeAttr<string S, list<AttrProperty> P> : Attr<S, P>;
 
 /// StringBool attribute.
 class StrBoolAttr<string S> : Attr<S, []>;
 
 /// Target-independent enum attributes.
 
 /// Alignment of parameter (5 bits) stored as log2 of alignment with +1 bias.
 /// 0 means unaligned (different from align(1)).
 def Alignment : IntAttr<"align", [ParamAttr, RetAttr]>;
 
 /// The result of the function is guaranteed to point to a number of bytes that
 /// we can determine if we know the value of the function's arguments.

 def NoInfsFPMath : StrBoolAttr<"no-infs-fp-math">;
 def NoNansFPMath : StrBoolAttr<"no-nans-fp-math">;
 def NoSignedZerosFPMath : StrBoolAttr<"no-signed-zeros-fp-math">;
 def UnsafeFPMath : StrBoolAttr<"unsafe-fp-math">;
 def NoJumpTables : StrBoolAttr<"no-jump-tables">;
 def NoInlineLineTables : StrBoolAttr<"no-inline-line-tables">;
 def ProfileSampleAccurate : StrBoolAttr<"profile-sample-accurate">;
 def UseSampleProfile : StrBoolAttr<"use-sample-profile">;
 
 class CompatRule<string F> {
   // The name of the function called to check the attribute of the caller and
   // callee and decide whether inlining should be allowed. The function's
   // signature must match "bool(const Function&, const Function &)", where the
   // first parameter is the reference to the caller and the second parameter is
   // the reference to the callee. It must return false if the attributes of the
   // caller and callee are incompatible, and true otherwise.
   string CompatFunc = F;
 }
 
-def : CompatRule<"isEqual<SanitizeAddressAttr>">;
-def : CompatRule<"isEqual<SanitizeThreadAttr>">;
-def : CompatRule<"isEqual<SanitizeMemoryAttr>">;
-def : CompatRule<"isEqual<SanitizeHWAddressAttr>">;
-def : CompatRule<"isEqual<SanitizeMemTagAttr>">;
-def : CompatRule<"isEqual<SafeStackAttr>">;
-def : CompatRule<"isEqual<ShadowCallStackAttr>">;
-def : CompatRule<"isEqual<UseSampleProfileAttr>">;
-def : CompatRule<"isEqual<NoProfileAttr>">;
+def : CompatRule<"isEqual<SanitizeAddress_Attr>">;
+def : CompatRule<"isEqual<SanitizeThread_Attr>">;
+def : CompatRule<"isEqual<SanitizeMemory_Attr>">;
+def : CompatRule<"isEqual<SanitizeHWAddress_Attr>">;
+def : CompatRule<"isEqual<SanitizeMemTag_Attr>">;
+def : CompatRule<"isEqual<SafeStack_Attr>">;
+def : CompatRule<"isEqual<ShadowCallStack_Attr>">;
+def : CompatRule<"isEqual<UseSampleProfile_Attr>">;
+def : CompatRule<"isEqual<NoProfile_Attr>">;
 
 class MergeRule<string F> {
   // The name of the function called to merge the attributes of the caller and
   // callee. The function's signature must match
   // "void(Function&, const Function &)", where the first parameter is the
   // reference to the caller and the second parameter is the reference to the
   // callee.
   string MergeFunc = F;
 }
 
-def : MergeRule<"setAND<LessPreciseFPMADAttr>">;
-def : MergeRule<"setAND<NoInfsFPMathAttr>">;
-def : MergeRule<"setAND<NoNansFPMathAttr>">;
-def : MergeRule<"setAND<NoSignedZerosFPMathAttr>">;
-def : MergeRule<"setAND<UnsafeFPMathAttr>">;
-def : MergeRule<"setOR<NoImplicitFloatAttr>">;
-def : MergeRule<"setOR<NoJumpTablesAttr>">;
-def : MergeRule<"setOR<ProfileSampleAccurateAttr>">;
-def : MergeRule<"setOR<SpeculativeLoadHardeningAttr>">;
+def : MergeRule<"setAND<LessPreciseFPMAD_Attr>">;
+def : MergeRule<"setAND<NoInfsFPMath_Attr>">;
+def : MergeRule<"setAND<NoNansFPMath_Attr>">;
+def : MergeRule<"setAND<NoSignedZerosFPMath_Attr>">;
+def : MergeRule<"setAND<UnsafeFPMath_Attr>">;
+def : MergeRule<"setOR<NoImplicitFloat_Attr>">;
+def : MergeRule<"setOR<NoJumpTables_Attr>">;
+def : MergeRule<"setOR<ProfileSampleAccurate_Attr>">;
+def : MergeRule<"setOR<SpeculativeLoadHardening_Attr>">;
 def : MergeRule<"adjustCallerSSPLevel">;
 def : MergeRule<"adjustCallerStackProbes">;
 def : MergeRule<"adjustCallerStackProbeSize">;
 def : MergeRule<"adjustMinLegalVectorWidth">;
 def : MergeRule<"adjustNullPointerValidAttr">;
-def : MergeRule<"setAND<MustProgressAttr>">;
+def : MergeRule<"setAND<MustProgress_Attr>">;

llvm/include/llvm/IR/Function.h

public:
   // TODO: remove non-AtIndex versions of these methods.
   /// adds the attribute to the list of attributes.
   void addAttributeAtIndex(unsigned i, Attribute Attr);
 
   /// Add function attributes to this function.
   void addFnAttr(Attribute::AttrKind Kind);
 
   /// Add function attributes to this function.
-  void addFnAttr(StringRef Kind, StringRef Val = StringRef());
+  void addFnAttr(AttributeKey Kind, StringRef Val = StringRef());
 
   /// Add function attributes to this function.
   void addFnAttr(Attribute Attr);
 
   /// Add function attributes to this function.
   void addFnAttrs(const AttrBuilder &Attrs);
 
   /// Add return value attributes to this function.
public:
 
   /// adds the attributes to the list of attributes for the given arg.
   void addParamAttrs(unsigned ArgNo, const AttrBuilder &Attrs);
 
   /// removes the attribute from the list of attributes.
   void removeAttributeAtIndex(unsigned i, Attribute::AttrKind Kind);
 
   /// removes the attribute from the list of attributes.
-  void removeAttributeAtIndex(unsigned i, StringRef Kind);
+  void removeAttributeAtIndex(unsigned i, AttributeKey Kind);
 
   /// Remove function attributes from this function.
   void removeFnAttr(Attribute::AttrKind Kind);
 
   /// Remove function attribute from this function.
-  void removeFnAttr(StringRef Kind);
+  void removeFnAttr(AttributeKey Kind);
 
   void removeFnAttrs(const AttrBuilder &Attrs);
 
   /// removes the attribute from the return value list of attributes.
   void removeRetAttr(Attribute::AttrKind Kind);
 
   /// removes the attribute from the return value list of attributes.
-  void removeRetAttr(StringRef Kind);
+  void removeRetAttr(AttributeKey Kind);
 
   /// removes the attributes from the return value list of attributes.
   void removeRetAttrs(const AttrBuilder &Attrs);
 
   /// removes the attribute from the list of attributes.
   void removeParamAttr(unsigned ArgNo, Attribute::AttrKind Kind);
 
   /// removes the attribute from the list of attributes.
-  void removeParamAttr(unsigned ArgNo, StringRef Kind);
+  void removeParamAttr(unsigned ArgNo, AttributeKey Kind);
 
   /// removes the attribute from the list of attributes.
   void removeParamAttrs(unsigned ArgNo, const AttrBuilder &Attrs);
 
   /// Return true if the function has the attribute.
   bool hasFnAttribute(Attribute::AttrKind Kind) const;
 
   /// Return true if the function has the attribute.
-  bool hasFnAttribute(StringRef Kind) const;
+  bool hasFnAttribute(AttributeKey Kind) const;
 
   /// check if an attribute is in the list of attributes for the return value.
   bool hasRetAttribute(Attribute::AttrKind Kind) const;
 
   /// check if an attributes is in the list of attributes.
   bool hasParamAttribute(unsigned ArgNo, Attribute::AttrKind Kind) const;
 
   /// gets the attribute from the list of attributes.
   Attribute getAttributeAtIndex(unsigned i, Attribute::AttrKind Kind) const;
 
   /// gets the attribute from the list of attributes.
-  Attribute getAttributeAtIndex(unsigned i, StringRef Kind) const;
+  Attribute getAttributeAtIndex(unsigned i, AttributeKey Kind) const;
 
   /// Return the attribute for the given attribute kind.
   Attribute getFnAttribute(Attribute::AttrKind Kind) const;
 
   /// Return the attribute for the given attribute kind.
-  Attribute getFnAttribute(StringRef Kind) const;
+  Attribute getFnAttribute(AttributeKey Kind) const;
 
   /// gets the specified attribute from the list of attributes.
   Attribute getParamAttribute(unsigned ArgNo, Attribute::AttrKind Kind) const;
 
   /// removes noundef and other attributes that imply undefined behavior if a
   /// `undef` or `poison` value is passed from the list of attributes.
   void removeParamUndefImplyingAttrs(unsigned ArgNo);
 
public:
   /// @param ArgNo AttributeList ArgNo, referring to an argument.
   uint64_t getParamDereferenceableOrNullBytes(unsigned ArgNo) const {
     return AttributeSets.getParamDereferenceableOrNullBytes(ArgNo);
   }
 
   /// A function will have the "coroutine.presplit" attribute if it's
   /// a coroutine and has not gone through full CoroSplit pass.
   bool isPresplitCoroutine() const {
-    return hasFnAttribute("coroutine.presplit");
+    return hasFnAttribute(CoroutinePresplitAttr);
   }
 
   /// Determine if the function does not access memory.
   bool doesNotAccessMemory() const {
     return hasFnAttribute(Attribute::ReadNone);
   }
   void setDoesNotAccessMemory() {
     addFnAttr(Attribute::ReadNone);

llvm/include/llvm/IR/GlobalVariable.h

public:
   void getDebugInfo(SmallVectorImpl<DIGlobalVariableExpression *> &GVs) const;
 
   /// Add attribute to this global.
   void addAttribute(Attribute::AttrKind Kind) {
     Attrs = Attrs.addAttribute(getContext(), Kind);
   }
 
   /// Add attribute to this global.
-  void addAttribute(StringRef Kind, StringRef Val = StringRef()) {
+  void addAttribute(AttributeKey Kind, StringRef Val = StringRef()) {
     Attrs = Attrs.addAttribute(getContext(), Kind, Val);
   }
 
   /// Return true if the attribute exists.
   bool hasAttribute(Attribute::AttrKind Kind) const {
     return Attrs.hasAttribute(Kind);
   }
 
   /// Return true if the attribute exists.
-  bool hasAttribute(StringRef Kind) const {
+  bool hasAttribute(AttributeKey Kind) const {
     return Attrs.hasAttribute(Kind);
   }
 
   /// Return true if any attributes exist.
   bool hasAttributes() const {
     return Attrs.hasAttributes();
   }
 
   /// Return the attribute object.
   Attribute getAttribute(Attribute::AttrKind Kind) const {
     return Attrs.getAttribute(Kind);
   }
 
   /// Return the attribute object.
-  Attribute getAttribute(StringRef Kind) const {
+  Attribute getAttribute(AttributeKey Kind) const {
     return Attrs.getAttribute(Kind);
   }
 
   /// Return the attribute set for this global
   AttributeSet getAttributes() const {
     return Attrs;
   }
 
public:
 
   /// Set attribute list for this global
   void setAttributes(AttributeSet A) {
     Attrs = A;
   }
 
   /// Check if section name is present
   bool hasImplicitSection() const {
-    return getAttributes().hasAttribute("bss-section") ||
-           getAttributes().hasAttribute("data-section") ||
-           getAttributes().hasAttribute("relro-section") ||
-           getAttributes().hasAttribute("rodata-section");
+    return getAttributes().hasAttribute(BssSectionAttr) ||
+           getAttributes().hasAttribute(DataSectionAttr) ||
+           getAttributes().hasAttribute(RelroSectionAttr) ||
+           getAttributes().hasAttribute(RodataSectionAttr);
   }
 
   // Methods for support type inquiry through isa, cast, and dyn_cast:
   static bool classof(const Value *V) {
     return V->getValueID() == Value::GlobalVariableVal;
   }
 };
 

llvm/include/llvm/IR/InstrTypes.h

     assert(Kind != Attribute::NoBuiltin &&
            "Use CallBase::isNoBuiltin() to check for Attribute::NoBuiltin");
     return hasFnAttrImpl(Kind);
   }
 
   /// Determine whether this call has the given attribute. If it does not
   /// then determine if the called function has the attribute, but only if
   /// the attribute is allowed for the call.
-  bool hasFnAttr(StringRef Kind) const { return hasFnAttrImpl(Kind); }
+  bool hasFnAttr(AttributeKey Kind) const { return hasFnAttrImpl(Kind); }
 
   // TODO: remove non-AtIndex versions of these methods.
   /// adds the attribute to the list of attributes.
   void addAttributeAtIndex(unsigned i, Attribute::AttrKind Kind) {
     Attrs = Attrs.addAttributeAtIndex(getContext(), i, Kind);
   }
 
   /// adds the attribute to the list of attributes.

   }
 
   /// removes the attribute from the list of attributes.
   void removeAttributeAtIndex(unsigned i, Attribute::AttrKind Kind) {
     Attrs = Attrs.removeAttributeAtIndex(getContext(), i, Kind);
   }
 
   /// removes the attribute from the list of attributes.
-  void removeAttributeAtIndex(unsigned i, StringRef Kind) {
+  void removeAttributeAtIndex(unsigned i, AttributeKey Kind) {
     Attrs = Attrs.removeAttributeAtIndex(getContext(), i, Kind);
   }
 
   /// Removes the attributes from the function
   void removeFnAttrs(const AttrBuilder &AttrsToRemove) {
     Attrs = Attrs.removeFnAttributes(getContext(), AttrsToRemove);
   }
 

 
   /// Removes the attribute from the given argument
   void removeParamAttr(unsigned ArgNo, Attribute::AttrKind Kind) {
     assert(ArgNo < arg_size() && "Out of bounds");
     Attrs = Attrs.removeParamAttribute(getContext(), ArgNo, Kind);
   }
 
   /// Removes the attribute from the given argument
-  void removeParamAttr(unsigned ArgNo, StringRef Kind) {
+  void removeParamAttr(unsigned ArgNo, AttributeKey Kind) {
     assert(ArgNo < arg_size() && "Out of bounds");
     Attrs = Attrs.removeParamAttribute(getContext(), ArgNo, Kind);
   }
 
   /// Removes the attributes from the given argument
   void removeParamAttrs(unsigned ArgNo, const AttrBuilder &AttrsToRemove) {
     Attrs = Attrs.removeParamAttributes(getContext(), ArgNo, AttrsToRemove);
   }
 
   /// adds the dereferenceable attribute to the list of attributes.
   void addDereferenceableParamAttr(unsigned i, uint64_t Bytes) {
     Attrs = Attrs.addDereferenceableParamAttr(getContext(), i, Bytes);
   }
 
   /// adds the dereferenceable attribute to the list of attributes.
   void addDereferenceableRetAttr(uint64_t Bytes) {
     Attrs = Attrs.addDereferenceableRetAttr(getContext(), Bytes);
   }
 
   /// Determine whether the return value has the given attribute.
   bool hasRetAttr(Attribute::AttrKind Kind) const {
     return hasRetAttrImpl(Kind);
   }
   /// Determine whether the return value has the given attribute.
-  bool hasRetAttr(StringRef Kind) const { return hasRetAttrImpl(Kind); }
+  bool hasRetAttr(AttributeKey Kind) const { return hasRetAttrImpl(Kind); }
 
   /// Determine whether the argument or parameter has the given attribute.
   bool paramHasAttr(unsigned ArgNo, Attribute::AttrKind Kind) const;
 
   /// Get the attribute of a given kind at a position.
   Attribute getAttributeAtIndex(unsigned i, Attribute::AttrKind Kind) const {
     return getAttributes().getAttributeAtIndex(i, Kind);
   }
 
   /// Get the attribute of a given kind at a position.
-  Attribute getAttributeAtIndex(unsigned i, StringRef Kind) const {
+  Attribute getAttributeAtIndex(unsigned i, AttributeKey Kind) const {
     return getAttributes().getAttributeAtIndex(i, Kind);
   }
 
   /// Get the attribute of a given kind for the function.
-  Attribute getFnAttr(StringRef Kind) const {
+  Attribute getFnAttr(AttributeKey Kind) const {
     return getAttributes().getFnAttr(Kind);
   }
 
   /// Get the attribute of a given kind for the function.
   Attribute getFnAttr(Attribute::AttrKind Kind) const {
     return getAttributes().getFnAttr(Kind);
   }
 
   /// Get the attribute of a given kind from a given arg
   Attribute getParamAttr(unsigned ArgNo, Attribute::AttrKind Kind) const {
     assert(ArgNo < arg_size() && "Out of bounds");
     return getAttributes().getParamAttr(ArgNo, Kind);
   }
 
   /// Get the attribute of a given kind from a given arg
-  Attribute getParamAttr(unsigned ArgNo, StringRef Kind) const {
+  Attribute getParamAttr(unsigned ArgNo, AttributeKey Kind) const {
     assert(ArgNo < arg_size() && "Out of bounds");
     return getAttributes().getParamAttr(ArgNo, Kind);
   }
 
   /// Return true if the data operand at index \p i has the attribute \p
   /// A.
   ///
   /// Data operands include call arguments and values used in operand bundles,

       if (!is_contained(IDs, ID))
         return true;
     }
     return false;
   }
 
   /// Is the function attribute S disallowed by some operand bundle on
   /// this operand bundle user?
-  bool isFnAttrDisallowedByOpBundle(StringRef S) const {
+  bool isFnAttrDisallowedByOpBundle(AttributeKey S) const {
     // Operand bundles only possibly disallow readnone, readonly and argmemonly
     // attributes.  All String attributes are fine.
     return false;
   }
 
   /// Is the function attribute A disallowed by some operand bundle on
   /// this operand bundle user?
   bool isFnAttrDisallowedByOpBundle(Attribute::AttrKind A) const {

     return Total;
   }
 
   /// @}
   // End of operand bundle API.
 
 private:
   bool hasFnAttrOnCalledFunction(Attribute::AttrKind Kind) const;
-  bool hasFnAttrOnCalledFunction(StringRef Kind) const;
+  bool hasFnAttrOnCalledFunction(AttributeKey Kind) const;
 
   template <typename AttrKind> bool hasFnAttrImpl(AttrKind Kind) const {
     if (Attrs.hasFnAttr(Kind))
       return true;
 
     // Operand bundles override attributes on the called function, but don't
     // override attributes directly present on the call instruction.
     if (isFnAttrDisallowedByOpBundle(Kind))
       return false;
 
     return hasFnAttrOnCalledFunction(Kind);
   }
 
   /// Determine whether the return value has the given attribute. Supports
-  /// Attribute::AttrKind and StringRef as \p AttrKind types.
+  /// Attribute::AttrKind and AttributeKey as \p AttrKind types.
   template <typename AttrKind> bool hasRetAttrImpl(AttrKind Kind) const {
     if (Attrs.hasRetAttr(Kind))
       return true;
 
     // Look at the callee, if available.
     if (const Function *F = getCalledFunction())
       return F->getAttributes().hasRetAttr(Kind);
     return false;

llvm/include/llvm/ProfileData/SampleProf.h

public:
 
   void setFunctionHash(uint64_t Hash) { FunctionHash = Hash; }
 
   uint64_t getFunctionHash() const { return FunctionHash; }
 
   /// Return the canonical name for a function, taking into account
   /// suffix elision policy attributes.
   static StringRef getCanonicalFnName(const Function &F) {
-    auto AttrName = "sample-profile-suffix-elision-policy";
-    auto Attr = F.getFnAttribute(AttrName).getValueAsString();
+    auto Attr = F.getFnAttribute(SampleProfileSuffixElisionPolicyAttr)
+                    .getValueAsString();
     return getCanonicalFnName(F.getName(), Attr);
   }
 
   /// Name suffixes which canonicalization should handle to avoid
   /// profile mismatch.
   static constexpr const char *LLVMSuffix = ".llvm.";
   static constexpr const char *PartSuffix = ".part.";
   static constexpr const char *UniqSuffix = ".__uniq.";

llvm/lib/Analysis/IVDescriptors.cpp

 
 bool RecurrenceDescriptor::isReductionPHI(PHINode *Phi, Loop *TheLoop,
                                           RecurrenceDescriptor &RedDes,
                                           DemandedBits *DB, AssumptionCache *AC,
                                           DominatorTree *DT) {
   BasicBlock *Header = TheLoop->getHeader();
   Function &F = *Header->getParent();
   FastMathFlags FMF;
-  FMF.setNoNaNs(
-      F.getFnAttribute("no-nans-fp-math").getValueAsBool());
+  FMF.setNoNaNs(F.getFnAttribute(NoNansFPMathAttr).getValueAsBool());
   FMF.setNoSignedZeros(
-      F.getFnAttribute("no-signed-zeros-fp-math").getValueAsBool());
+      F.getFnAttribute(NoSignedZerosFPMathAttr).getValueAsBool());
 
   if (AddReductionVar(Phi, RecurKind::Add, TheLoop, FMF, RedDes, DB, AC, DT)) {
     LLVM_DEBUG(dbgs() << "Found an ADD reduction PHI." << *Phi << "\n");
     return true;
   }
   if (AddReductionVar(Phi, RecurKind::Mul, TheLoop, FMF, RedDes, DB, AC, DT)) {
     LLVM_DEBUG(dbgs() << "Found a MUL reduction PHI." << *Phi << "\n");
     return true;

llvm/lib/Analysis/InlineAdvisor.cpp

std::string llvm::inlineCostStr(const InlineCost &IC) {
   Remark << IC;
   return Remark.str();
 }
 
 void llvm::setInlineRemark(CallBase &CB, StringRef Message) {
   if (!InlineRemarkAttribute)
     return;
 
-  Attribute Attr = Attribute::get(CB.getContext(), "inline-remark", Message);
+  Attribute Attr = Attribute::get(CB.getContext(), InlineRemarkAttr, Message);
   CB.addFnAttr(Attr);
 }
 
 /// Return the cost only if the inliner should attempt to inline at the given
 /// CallSite. If we return the cost, we will emit an optimisation remark later
 /// using that cost, so we won't do so from this function. Return None if
 /// inlining should not be attempted.
 Optional<InlineCost>

llvm/lib/Analysis/InlineCost.cpp

static cl::opt<bool> DisableGEPConstOperand(
     cl::desc("Disables evaluation of GetElementPtr with constant operands"));
 
 namespace {
 class InlineCostCallAnalyzer;
 
 /// This function behaves more like CallBase::hasFnAttr: when it looks for the
 /// requested attribute, it check both the call instruction and the called
 /// function (if it's available and operand bundles don't prohibit that).
-Attribute getFnAttr(CallBase &CB, StringRef AttrKind) {
+Attribute getFnAttr(CallBase &CB, AttributeKey AttrKind) {
   Attribute CallAttr = CB.getFnAttr(AttrKind);
   if (CallAttr.isValid())
     return CallAttr;
 
   // Operand bundles override attributes on the called function, but don't
   // override attributes directly present on the call instruction.
   if (!CB.isFnAttrDisallowedByOpBundle(AttrKind))
     if (const Function *F = CB.getCalledFunction())
       return F->getFnAttribute(AttrKind);
 
   return {};
 }
 
-Optional<int> getStringFnAttrAsInt(CallBase &CB, StringRef AttrKind) {
+Optional<int> getStringFnAttrAsInt(CallBase &CB, AttributeKey AttrKind) {
   Attribute Attr = getFnAttr(CB, AttrKind);
   int AttrValue;
   if (Attr.getValueAsString().getAsInteger(10, AttrValue))
     return None;
   return AttrValue;
 }
 
 // This struct is used to store information about inline cost of a
class InlineCostCallAnalyzer final : public CallAnalyzer {
 
   void onDisableLoadElimination() override {
     addCost(LoadEliminationCost);
     LoadEliminationCost = 0;
   }
 
   bool onCallBaseVisitStart(CallBase &Call) override {
     if (Optional<int> AttrCallThresholdBonus =
-            getStringFnAttrAsInt(Call, "call-threshold-bonus"))
+            getStringFnAttrAsInt(Call, CallThresholdBonus))
       Threshold += *AttrCallThresholdBonus;
 
     if (Optional<int> AttrCallCost =
-            getStringFnAttrAsInt(Call, "call-inline-cost")) {
+            getStringFnAttrAsInt(Call, CallInlineCost)) {
       addCost(*AttrCallCost);
       // Prevent further processing of the call since we want to override its
       // inline cost, not just add to it.
       return false;
     }
     return true;
   }
 
InlineResult finalizeAnalysis() override {
     // subtract the excess bonus, if any, from the Threshold before
     // comparing against Cost.
     if (NumVectorInstructions <= NumInstructions / 10)
       Threshold -= VectorBonus;
     else if (NumVectorInstructions <= NumInstructions / 2)
       Threshold -= VectorBonus / 2;
 
     if (Optional<int> AttrCost =
-            getStringFnAttrAsInt(CandidateCall, "function-inline-cost"))
+            getStringFnAttrAsInt(CandidateCall, FunctionInlineCost))
       Cost = *AttrCost;
 
     if (Optional<int> AttrThreshold =
-            getStringFnAttrAsInt(CandidateCall, "function-inline-threshold"))
+            getStringFnAttrAsInt(CandidateCall, FunctionInlineThreshold))
       Threshold = *AttrThreshold;
 
     if (auto Result = costBenefitAnalysis()) {
       DecidedByCostBenefit = true;
       if (Result.getValue())
         return InlineResult::success();
       else
         return InlineResult::failure("Cost over threshold.");

llvm/lib/Analysis/VectorUtils.cpp

   for (unsigned I = 0; I < numArgs; ++I)
     Out << "v";
   Out << "_" << ScalarName << "(" << VectorName << ")";
   return std::string(Out.str());
 }
 
 void VFABI::getVectorVariantNames(
     const CallInst &CI, SmallVectorImpl<std::string> &VariantMappings) {
-  const StringRef S = CI.getFnAttr(VFABI::MappingsAttrName).getValueAsString();
+  const StringRef S =
+      CI.getFnAttr(VectorFunctionAbiVariantAttr).getValueAsString();
   if (S.empty())
     return;
 
   SmallVector<StringRef, 8> ListAttr;
   S.split(ListAttr, ",");
 
   for (auto &S : SetVector<StringRef>(ListAttr.begin(), ListAttr.end())) {
 #ifndef NDEBUG

llvm/lib/AsmParser/LLParser.cpp

 ///   := StringConstant
 ///   := StringConstant '=' StringConstant
 bool LLParser::parseStringAttribute(AttrBuilder &B) {
   std::string Attr = Lex.getStrVal();
   Lex.Lex();
   std::string Val;
   if (EatIfPresent(lltok::equal) && parseStringConstant(Val))
     return true;
-  B.addAttribute(Attr, Val);
+  B.addAttribute(AttributeKey::Create(Attr), Val);
   return false;
 }
 
 /// Parse a potentially empty list of parameter or return attributes.
 bool LLParser::parseOptionalParamOrReturnAttrs(AttrBuilder &B, bool IsParam) {
   bool HaveError = false;
 
   B.clear();

llvm/lib/Bitcode/Reader/BitcodeReader.cpp

           if (HasValue) {
             // Has a value associated with it.
             ++i; // Skip the '0' that terminates the "kind" string.
             while (Record[i] != 0 && i != e)
               ValStr += Record[i++];
             assert(Record[i] == 0 && "Value string not null terminated");
           }
 
-          B.addAttribute(KindStr.str(), ValStr.str());
+          B.addAttribute(AttributeKey::Create(KindStr), ValStr.str());
         } else {
           assert((Record[i] == 5 || Record[i] == 6) &&
                  "Invalid attribute group entry");
           bool HasType = Record[i] == 6;
           Attribute::AttrKind Kind;
           if (Error Err = parseAttrKind(Record[++i], &Kind))
             return Err;
           if (!Attribute::isTypeAttrKind(Kind))

llvm/lib/Bitcode/Writer/BitcodeWriter.cpp

for (Attribute Attr : AS) {
       if (Attr.isEnumAttribute()) {
         Record.push_back(0);
         Record.push_back(getAttrKindEncoding(Attr.getKindAsEnum()));
       } else if (Attr.isIntAttribute()) {
         Record.push_back(1);
         Record.push_back(getAttrKindEncoding(Attr.getKindAsEnum()));
         Record.push_back(Attr.getValueAsInt());
       } else if (Attr.isStringAttribute()) {
-        StringRef Kind = Attr.getKindAsString();
+        StringRef Kind = Attr.getKindAsKey().value();
         StringRef Val = Attr.getValueAsString();
 
         Record.push_back(Val.empty() ? 3 : 4);
         Record.append(Kind.begin(), Kind.end());
         Record.push_back(0);
         if (!Val.empty()) {
           Record.append(Val.begin(), Val.end());
           Record.push_back(0);

llvm/lib/CodeGen/AsmPrinter/AsmPrinter.cpp

if (MAI->hasSubsectionsViaSymbols()) {
       emitGlobalConstant(F.getParent()->getDataLayout(), F.getPrefixData());
     }
   }
 
   // Emit M NOPs for -fpatchable-function-entry=N,M where M>0. We arbitrarily
   // place prefix data before NOPs.
   unsigned PatchableFunctionPrefix = 0;
   unsigned PatchableFunctionEntry = 0;
-  (void)F.getFnAttribute("patchable-function-prefix")
+  (void)F.getFnAttribute(PatchableFunctionPrefixAttr)
       .getValueAsString()
       .getAsInteger(10, PatchableFunctionPrefix);
-  (void)F.getFnAttribute("patchable-function-entry")
+  (void)F.getFnAttribute(PatchableFunctionEntryAttr)
       .getValueAsString()
       .getAsInteger(10, PatchableFunctionEntry);
   if (PatchableFunctionPrefix) {
     CurrentPatchableFunctionEntrySym =
         OutContext.createLinkerPrivateTempSymbol();
     OutStreamer->emitLabel(CurrentPatchableFunctionEntrySym);
     emitNops(PatchableFunctionPrefix);
   } else if (PatchableFunctionEntry) {
void AsmPrinter::SetupMachineFunction(MachineFunction &MF) {
   }
 
   CurrentFnSymForSize = CurrentFnSym;
   CurrentFnBegin = nullptr;
   CurrentSectionBeginSym = nullptr;
   MBBSectionRanges.clear();
   MBBSectionExceptionSyms.clear();
   bool NeedsLocalForSize = MAI->needsLocalForSize();
-  if (F.hasFnAttribute("patchable-function-entry") ||
-      F.hasFnAttribute("function-instrument") ||
-      F.hasFnAttribute("xray-instruction-threshold") ||
+  if (F.hasFnAttribute(PatchableFunctionEntryAttr) ||
+      F.hasFnAttribute(FunctionInstrumentAttr) ||
+      F.hasFnAttribute(XrayInstructionThresholdAttr) ||
       needFuncLabelsForEHOrDebugInfo(MF) || NeedsLocalForSize ||
       MF.getTarget().Options.EmitStackSizeSection || MF.hasBBLabels()) {
     CurrentFnBegin = createTempSymbol("func_begin");
     if (NeedsLocalForSize)
       CurrentFnSymForSize = CurrentFnBegin;
   }
 
   ORE = &getAnalysis<MachineOptimizationRemarkEmitterPass>().getORE();
if (FnSledIndex) {
     OutStreamer->SwitchSection(PrevSection);
   }
   Sleds.clear();
 }
 
 void AsmPrinter::recordSled(MCSymbol *Sled, const MachineInstr &MI,
                             SledKind Kind, uint8_t Version) {
   const Function &F = MI.getMF()->getFunction();
-  auto Attr = F.getFnAttribute("function-instrument");
-  bool LogArgs = F.hasFnAttribute("xray-log-args");
+  auto Attr = F.getFnAttribute(FunctionInstrumentAttr);
+  bool LogArgs = F.hasFnAttribute(XrayLogArgsAttr);
   bool AlwaysInstrument =
     Attr.isStringAttribute() && Attr.getValueAsString() == "xray-always";
   if (Kind == SledKind::FUNCTION_ENTER && LogArgs)
     Kind = SledKind::LOG_ARGS_ENTER;
   Sleds.emplace_back(XRayFunctionEntry{Sled, CurrentFnSym, Kind,
                                        AlwaysInstrument, &F, Version});
 }
 
 void AsmPrinter::emitPatchableFunctionEntries() {
   const Function &F = MF->getFunction();
   unsigned PatchableFunctionPrefix = 0, PatchableFunctionEntry = 0;
-  (void)F.getFnAttribute("patchable-function-prefix")
+  (void)F.getFnAttribute(PatchableFunctionPrefixAttr)
       .getValueAsString()
       .getAsInteger(10, PatchableFunctionPrefix);
-  (void)F.getFnAttribute("patchable-function-entry")
+  (void)F.getFnAttribute(PatchableFunctionEntryAttr)
       .getValueAsString()
       .getAsInteger(10, PatchableFunctionEntry);
   if (!PatchableFunctionPrefix && !PatchableFunctionEntry)
     return;
   const unsigned PointerSize = getPointerSize();
   if (TM.getTargetTriple().isOSBinFormatELF()) {
     auto Flags = ELF::SHF_WRITE | ELF::SHF_ALLOC;
     const MCSymbolELF *LinkedToSym = nullptr;

llvm/lib/CodeGen/AsmPrinter/WinException.cpp

 WinException::~WinException() {}
 
 /// endModule - Emit all exception information that should come after the
 /// content.
 void WinException::endModule() {
   auto &OS = *Asm->OutStreamer;
   const Module *M = MMI->getModule();
   for (const Function &F : *M)
-    if (F.hasFnAttribute("safeseh"))
+    if (F.hasFnAttribute(SafesehAttr))
       OS.EmitCOFFSafeSEH(Asm->getSymbol(&F));
 
   if (M->getModuleFlag("ehcontguard") && !EHContTargets.empty()) {
     // Emit the symbol index of each ehcont target.
     OS.SwitchSection(Asm->OutContext.getObjectFileInfo()->getGEHContSection());
     for (const MCSymbol *S : EHContTargets) {
       OS.EmitCOFFSymbolIndex(S);
     }

llvm/lib/CodeGen/CommandFlags.cpp

std::vector<std::string> codegen::getFeatureList() {
   }
 
   for (auto const &MAttr : getMAttrs())
     Features.AddFeature(MAttr);
 
   return Features.getFeatures();
 }
 
-void codegen::renderBoolStringAttr(AttrBuilder &B, StringRef Name, bool Val) {
+void codegen::renderBoolStringAttr(AttrBuilder &B, AttributeKey Name,
+                                   bool Val) {
   B.addAttribute(Name, Val ? "true" : "false");
 }
 
-#define HANDLE_BOOL_ATTR(CL, AttrName)                                         \
+#define HANDLE_BOOL_ATTR(CL, Attr)                                             \
   do {                                                                         \
-    if (CL->getNumOccurrences() > 0 && !F.hasFnAttribute(AttrName))            \
-      renderBoolStringAttr(NewAttrs, AttrName, *CL);                           \
+    if (CL->getNumOccurrences() > 0 && !F.hasFnAttribute(Attr))                \
+      renderBoolStringAttr(NewAttrs, Attr, *CL);                               \
   } while (0)
 
 /// Set function attributes of function \p F based on CPU, Features, and command
 /// line flags.
 void codegen::setFunctionAttributes(StringRef CPU, StringRef Features,
                                     Function &F) {
   auto &Ctx = F.getContext();
   AttributeList Attrs = F.getAttributes();
   AttrBuilder NewAttrs;
 
-  if (!CPU.empty() && !F.hasFnAttribute("target-cpu"))
-    NewAttrs.addAttribute("target-cpu", CPU);
+  if (!CPU.empty() && !F.hasFnAttribute(TargetCPUAttr))
+    NewAttrs.addAttribute(TargetCPUAttr, CPU);
   if (!Features.empty()) {
     // Append the command line features to any that are already on the function.
     StringRef OldFeatures =
-        F.getFnAttribute("target-features").getValueAsString();
+        F.getFnAttribute(TargetFeaturesAttr).getValueAsString();
     if (OldFeatures.empty())
-      NewAttrs.addAttribute("target-features", Features);
+      NewAttrs.addAttribute(TargetFeaturesAttr, Features);
     else {
       SmallString<256> Appended(OldFeatures);
       Appended.push_back(',');
       Appended.append(Features);
-      NewAttrs.addAttribute("target-features", Appended);
+      NewAttrs.addAttribute(TargetFeaturesAttr, Appended);
     }
   }
   if (FramePointerUsageView->getNumOccurrences() > 0 &&
-      !F.hasFnAttribute("frame-pointer")) {
+      !F.hasFnAttribute(FramePointerAttr)) {
     if (getFramePointerUsage() == FramePointerKind::All)
-      NewAttrs.addAttribute("frame-pointer", "all");
+      NewAttrs.addAttribute(FramePointerAttr, "all");
     else if (getFramePointerUsage() == FramePointerKind::NonLeaf)
-      NewAttrs.addAttribute("frame-pointer", "non-leaf");
+      NewAttrs.addAttribute(FramePointerAttr, "non-leaf");
     else if (getFramePointerUsage() == FramePointerKind::None)
-      NewAttrs.addAttribute("frame-pointer", "none");
+      NewAttrs.addAttribute(FramePointerAttr, "none");
   }
   if (DisableTailCallsView->getNumOccurrences() > 0)
-    NewAttrs.addAttribute("disable-tail-calls",
+    NewAttrs.addAttribute(DisableTailCallsAttr,
                           toStringRef(getDisableTailCalls()));
   if (getStackRealign())
-    NewAttrs.addAttribute("stackrealign");
+    NewAttrs.addAttribute(StackrealignAttr);
 
-  HANDLE_BOOL_ATTR(EnableUnsafeFPMathView, "unsafe-fp-math");
-  HANDLE_BOOL_ATTR(EnableNoInfsFPMathView, "no-infs-fp-math");
-  HANDLE_BOOL_ATTR(EnableNoNaNsFPMathView, "no-nans-fp-math");
-  HANDLE_BOOL_ATTR(EnableNoSignedZerosFPMathView, "no-signed-zeros-fp-math");
+  HANDLE_BOOL_ATTR(EnableUnsafeFPMathView, UnsafeFPMathAttr);
+  HANDLE_BOOL_ATTR(EnableNoInfsFPMathView, NoInfsFPMathAttr);
+  HANDLE_BOOL_ATTR(EnableNoNaNsFPMathView, NoNansFPMathAttr);
+  HANDLE_BOOL_ATTR(EnableNoSignedZerosFPMathView, NoSignedZerosFPMathAttr);
 
   if (DenormalFPMathView->getNumOccurrences() > 0 &&
-      !F.hasFnAttribute("denormal-fp-math")) {
+      !F.hasFnAttribute(DenormalFPMathAttr)) {
     DenormalMode::DenormalModeKind DenormKind = getDenormalFPMath();
 
     // FIXME: Command line flag should expose separate input/output modes.
-    NewAttrs.addAttribute("denormal-fp-math",
+    NewAttrs.addAttribute(DenormalFPMathAttr,
                           DenormalMode(DenormKind, DenormKind).str());
   }
 
   if (DenormalFP32MathView->getNumOccurrences() > 0 &&
-      !F.hasFnAttribute("denormal-fp-math-f32")) {
+      !F.hasFnAttribute(DenormalFPMathf32Attr)) {
     // FIXME: Command line flag should expose separate input/output modes.
     DenormalMode::DenormalModeKind DenormKind = getDenormalFP32Math();
 
-    NewAttrs.addAttribute(
-      "denormal-fp-math-f32",
+    NewAttrs.addAttribute(DenormalFPMathf32Attr,
                           DenormalMode(DenormKind, DenormKind).str());
   }
 
   if (TrapFuncNameView->getNumOccurrences() > 0)
     for (auto &B : F)
       for (auto &I : B)
         if (auto *Call = dyn_cast<CallInst>(&I))
           if (const auto *F = Call->getCalledFunction())
             if (F->getIntrinsicID() == Intrinsic::debugtrap ||
                 F->getIntrinsicID() == Intrinsic::trap)
               Call->addFnAttr(
-                  Attribute::get(Ctx, "trap-func-name", getTrapFuncName()));
+                  Attribute::get(Ctx, TrapFuncNameAttr, getTrapFuncName()));
 
   // Let NewAttrs override Attrs.
   F.setAttributes(Attrs.addFnAttributes(Ctx, NewAttrs));
 }
 
 /// Set function attributes of functions in Module M based on CPU,
 /// Features, and command line flags.
 void codegen::setFunctionAttributes(StringRef CPU, StringRef Features,
                                     Module &M) {
   for (Function &F : M)
     setFunctionAttributes(CPU, Features, F);
 }

llvm/lib/CodeGen/FEntryInserter.cpp

struct FEntryInserter : public MachineFunctionPass {
   }
 
   bool runOnMachineFunction(MachineFunction &F) override;
 };
 }
 
 bool FEntryInserter::runOnMachineFunction(MachineFunction &MF) {
   const std::string FEntryName = std::string(
-      MF.getFunction().getFnAttribute("fentry-call").getValueAsString());
+      MF.getFunction().getFnAttribute(FentryCallAttr).getValueAsString());
   if (FEntryName != "true")
     return false;
 
   auto &FirstMBB = *MF.begin();
   auto *TII = MF.getSubtarget().getInstrInfo();
   BuildMI(FirstMBB, FirstMBB.begin(), DebugLoc(),
           TII->get(TargetOpcode::FENTRY_CALL));
   return true;
 }
 
 char FEntryInserter::ID = 0;
 char &llvm::FEntryInserterID = FEntryInserter::ID;
 INITIALIZE_PASS(FEntryInserter, "fentry-insert", "Insert fentry calls", false,
                 false)

llvm/lib/CodeGen/GlobalISel/CallLowering.cpp

bool CallLowering::lowerCall(MachineIRBuilder &MIRBuilder, const CallBase &CB,
                              Register SwiftErrorVReg,
                              std::function<unsigned()> GetCalleeReg) const {
   CallLoweringInfo Info;
   const DataLayout &DL = MIRBuilder.getDataLayout();
   MachineFunction &MF = MIRBuilder.getMF();
   bool CanBeTailCalled = CB.isTailCall() &&
                          isInTailCallPosition(CB, MF.getTarget()) &&
                          (MF.getFunction()
-                              .getFnAttribute("disable-tail-calls")
+                              .getFnAttribute(DisableTailCallsAttr)
                               .getValueAsString() != "true");
 
   CallingConv::ID CallConv = CB.getCallingConv();
   Type *RetTy = CB.getType();
   bool IsVarArg = CB.getFunctionType()->isVarArg();
 
   SmallVector<BaseArgInfo, 4> SplitArgs;
   getReturnInfo(CallConv, RetTy, CB.getAttributes(), SplitArgs, DL);

llvm/lib/CodeGen/GlobalISel/IRTranslator.cpp

                           MachineInstr::copyFlagsFromInstruction(CI));
 
     return true;
   }
   case Intrinsic::trap:
   case Intrinsic::debugtrap:
   case Intrinsic::ubsantrap: {
     StringRef TrapFuncName =
-        CI.getAttributes().getFnAttr("trap-func-name").getValueAsString();
+        CI.getAttributes().getFnAttr(TrapFuncNameAttr).getValueAsString();
     if (TrapFuncName.empty())
       break; // Use the default handling.
     CallLowering::CallLoweringInfo Info;
     if (ID == Intrinsic::ubsantrap) {
       Info.OrigArgs.push_back({getOrCreateVRegs(*CI.getArgOperand(0)),
                                CI.getArgOperand(0)->getType(), 0});
     }
     Info.Callee = MachineOperand::CreateES(TrapFuncName.data());

llvm/lib/CodeGen/MachineFunction.cpp

if (STI->getRegisterInfo())
     RegInfo = new (Allocator) MachineRegisterInfo(this);
   else
     RegInfo = nullptr;
 
   MFInfo = nullptr;
   // We can realign the stack if the target supports it and the user hasn't
   // explicitly asked us not to.
   bool CanRealignSP = STI->getFrameLowering()->isStackRealignable() &&
-                      !F.hasFnAttribute("no-realign-stack");
+                      !F.hasFnAttribute(NoRealignStack);
   FrameInfo = new (Allocator) MachineFrameInfo(
       getFnStackAlignment(STI, F), /*StackRealignable=*/CanRealignSP,
       /*ForcedRealign=*/CanRealignSP &&
           F.hasFnAttribute(Attribute::StackAlignment));
 
   if (F.hasFnAttribute(Attribute::StackAlignment))
     FrameInfo->ensureMaxAlignment(*F.getFnStackAlign());
 

 }
 
 DenormalMode MachineFunction::getDenormalMode(const fltSemantics &FPType) const {
   return F.getDenormalMode(FPType);
 }
 
 /// Should we be emitting segmented stack stuff for the function
 bool MachineFunction::shouldSplitStack() const {
-  return getFunction().hasFnAttribute("split-stack");
+  return getFunction().hasFnAttribute(SplitStackAttr);
 }
 
 LLVM_NODISCARD unsigned
 MachineFunction::addFrameInst(const MCCFIInstruction &Inst) {
   FrameInstructions.push_back(Inst);
   return FrameInstructions.size() - 1;
 }
 

llvm/lib/CodeGen/MachineFunctionSplitter.cpp

bool MachineFunctionSplitter::runOnMachineFunction(MachineFunction &MF) {
   if (!MF.getFunction().hasProfileData())
     return false;
 
   // TODO: We don't split functions where a section attribute has been set
   // since the split part may not be placed in a contiguous region. It may also
   // be more beneficial to augment the linker to ensure contiguous layout of
   // split functions within the same section as specified by the attribute.
   if (MF.getFunction().hasSection() ||
-      MF.getFunction().hasFnAttribute("implicit-section-name"))
+      MF.getFunction().hasFnAttribute(ImplicitSectionNameAttr))
     return false;
 
   // We don't want to proceed further for cold functions
   // or functions of unknown hotness. Lukewarm functions have no prefix.
   Optional<StringRef> SectionPrefix = MF.getFunction().getSectionPrefix();
   if (SectionPrefix.hasValue() &&
       (SectionPrefix.getValue().equals("unlikely") ||
        SectionPrefix.getValue().equals("unknown"))) {

llvm/lib/CodeGen/MachineOutliner.cpp

MachineFunction *MachineOutliner::createOutlinedFunction(
   F->addFnAttr(Attribute::MinSize);
 
   // Include target features from an arbitrary candidate for the outlined
   // function. This makes sure the outlined function knows what kinds of
   // instructions are going into it. This is fine, since all parent functions
   // must necessarily support the instructions that are in the outlined region.
   Candidate &FirstCand = OF.Candidates.front();
   const Function &ParentFn = FirstCand.getMF()->getFunction();
-  if (ParentFn.hasFnAttribute("target-features"))
-    F->addFnAttr(ParentFn.getFnAttribute("target-features"));
+  if (ParentFn.hasFnAttribute(TargetFeaturesAttr))
+    F->addFnAttr(ParentFn.getFnAttribute(TargetFeaturesAttr));
 
   // Set nounwind, so we don't generate eh_frame.
   if (llvm::all_of(OF.Candidates, [](const outliner::Candidate &C) {
         return C.getMF()->getFunction().hasFnAttribute(Attribute::NoUnwind);
       }))
     F->addFnAttr(Attribute::NoUnwind);
 
   BasicBlock *EntryBB = BasicBlock::Create(C, "entry", F);

llvm/lib/CodeGen/PatchableFunction.cpp

static bool doesNotGeneratecode(const MachineInstr &MI) {
   case TargetOpcode::GC_LABEL:
   case TargetOpcode::DBG_VALUE:
   case TargetOpcode::DBG_LABEL:
     return true;
   }
 }
 
 bool PatchableFunction::runOnMachineFunction(MachineFunction &MF) {
-  if (MF.getFunction().hasFnAttribute("patchable-function-entry")) {
+  if (MF.getFunction().hasFnAttribute(PatchableFunctionEntryAttr)) {
     MachineBasicBlock &FirstMBB = *MF.begin();
     const TargetInstrInfo *TII = MF.getSubtarget().getInstrInfo();
     // The initial .loc covers PATCHABLE_FUNCTION_ENTER.
     BuildMI(FirstMBB, FirstMBB.begin(), DebugLoc(),
             TII->get(TargetOpcode::PATCHABLE_FUNCTION_ENTER));
     return true;
   }
 
-  if (!MF.getFunction().hasFnAttribute("patchable-function"))
+  if (!MF.getFunction().hasFnAttribute(PatchableFunctionAttr))
     return false;
 
 #ifndef NDEBUG
-  Attribute PatchAttr = MF.getFunction().getFnAttribute("patchable-function");
+  Attribute PatchAttr = MF.getFunction().getFnAttribute(PatchableFunctionAttr);
   StringRef PatchType = PatchAttr.getValueAsString();
   assert(PatchType == "prologue-short-redirect" && "Only possibility today!");
 #endif
 
   auto &FirstMBB = *MF.begin();
   MachineBasicBlock::iterator FirstActualI = FirstMBB.begin();
   for (; doesNotGeneratecode(*FirstActualI); ++FirstActualI)
     assert(FirstActualI != FirstMBB.end());

llvm/lib/CodeGen/PrologEpilogInserter.cpp

bool PEI::runOnMachineFunction(MachineFunction &MF) {
   if (TRI->requiresRegisterScavenging(MF) && FrameIndexVirtualScavenging)
     scavengeFrameVirtualRegs(MF, *RS);
 
   // Warn on stack size when we exceeds the given limit.
   MachineFrameInfo &MFI = MF.getFrameInfo();
   uint64_t StackSize = MFI.getStackSize();
 
   unsigned Threshold = UINT_MAX;
-  if (MF.getFunction().hasFnAttribute("warn-stack-size")) {
+  if (MF.getFunction().hasFnAttribute(WarnStackSizeAttr)) {
     bool Failed = MF.getFunction()
-                      .getFnAttribute("warn-stack-size")
+                      .getFnAttribute(WarnStackSizeAttr)
                       .getValueAsString()
                       .getAsInteger(10, Threshold);
     // Verifier should have caught this.
     assert(!Failed && "Invalid warn-stack-size fn attr value");
     (void)Failed;
   }
   if (StackSize > Threshold) {
     DiagnosticInfoStackSize DiagStackSize(F, StackSize, Threshold, DS_Warning);

llvm/lib/CodeGen/SelectionDAG/DAGCombiner.cpp

 }
 
 static SDValue foldFPToIntToFP(SDNode *N, SelectionDAG &DAG,
                                const TargetLowering &TLI) {
   // This optimization is guarded by a function attribute because it may produce
   // unexpected results. Ie, programs may be relying on the platform-specific
   // undefined behavior when the float-to-int conversion overflows.
   const Function &F = DAG.getMachineFunction().getFunction();
-  Attribute StrictOverflow = F.getFnAttribute("strict-float-cast-overflow");
+  Attribute StrictOverflow = F.getFnAttribute(StrictFloatCastOverflowAttr);
   if (StrictOverflow.getValueAsString().equals("false"))
     return SDValue();
 
   // We only do this if the target has legal ftrunc. Otherwise, we'd likely be
   // replacing casts with a libcall. We also must be allowed to ignore -0.0
   // because FTRUNC will return -0.0 for (-1.0, -0.0), but using integer
   // conversions would return +0.0.
   // FIXME: We should be able to use node-level FMF here.

llvm/lib/CodeGen/SelectionDAG/FastISel.cpp

     Args.push_back(Entry);
   }
 
   // Check if target-independent constraints permit a tail call here.
   // Target-dependent constraints are checked within fastLowerCall.
   bool IsTailCall = CI->isTailCall();
   if (IsTailCall && !isInTailCallPosition(*CI, TM))
     IsTailCall = false;
-  if (IsTailCall && MF->getFunction()
-                            .getFnAttribute("disable-tail-calls")
-                            .getValueAsBool())
+  if (IsTailCall &&
+      MF->getFunction().getFnAttribute(DisableTailCallsAttr).getValueAsBool())
     IsTailCall = false;
 
   CallLoweringInfo CLI;
   CLI.setCallee(RetTy, FuncTy, CI->getCalledOperand(), std::move(Args), *CI)
       .setTailCall(IsTailCall);
 
   diagnoseDontCall(*CI);
 

     // may be translated directly to target instructions.
     if (F && !F->hasLocalLinkage() && F->hasName() &&
         LibInfo->getLibFunc(F->getName(), Func) &&
         LibInfo->hasOptimizedCodeGen(Func))
       return false;
 
     // Don't handle Intrinsic::trap if a trap function is specified.
     if (F && F->getIntrinsicID() == Intrinsic::trap &&
-        Call->hasFnAttr("trap-func-name"))
+        Call->hasFnAttr(TrapFuncNameAttr))
       return false;
   }
 
   // First, try doing target-independent selection.
   if (!SkipTargetIndependentISel) {
     if (selectOperator(I, I->getOpcode())) {
       ++NumFastIselSuccessIndependent;
       DbgLoc = DebugLoc();

llvm/lib/CodeGen/SelectionDAG/SelectionDAGBuilder.cpp

     // Just replace __builtin_expect(exp, c) with EXP.
     setValue(&I, getValue(I.getArgOperand(0)));
     return;
 
   case Intrinsic::ubsantrap:
   case Intrinsic::debugtrap:
   case Intrinsic::trap: {
     StringRef TrapFuncName =
-        I.getAttributes().getFnAttr("trap-func-name").getValueAsString();
+        I.getAttributes().getFnAttr(TrapFuncNameAttr).getValueAsString();
     if (TrapFuncName.empty()) {
       switch (Intrinsic) {
       case Intrinsic::trap:
         DAG.setRoot(DAG.getNode(ISD::TRAP, sdl, MVT::Other, getRoot()));
         break;
       case Intrinsic::debugtrap:
         DAG.setRoot(DAG.getNode(ISD::DEBUGTRAP, sdl, MVT::Other, getRoot()));
         break;

 
   const Value *SwiftErrorVal = nullptr;
   const TargetLowering &TLI = DAG.getTargetLoweringInfo();
 
   if (isTailCall) {
     // Avoid emitting tail calls in functions with the disable-tail-calls
     // attribute.
     auto *Caller = CB.getParent()->getParent();
-    if (Caller->getFnAttribute("disable-tail-calls").getValueAsString() ==
-        "true" && !isMustTailCall)
+    if (Caller->getFnAttribute(DisableTailCallsAttr).getValueAsString() ==
+            "true" &&
+        !isMustTailCall)
       isTailCall = false;
 
     // We can't tail call inside a function with a swifterror argument. Lowering
     // does not support this yet. It would have to move into the swifterror
     // register before the call.
     if (TLI.supportSwiftError() &&
         Caller->getAttributes().hasAttrSomewhere(Attribute::SwiftError))
       isTailCall = false;

llvm/lib/CodeGen/SelectionDAG/TargetLowering.cpp

 
 /// Check whether a given call node is in tail position within its function. If
 /// so, it sets Chain to the input chain of the tail call.
 bool TargetLowering::isInTailCallPosition(SelectionDAG &DAG, SDNode *Node,
                                           SDValue &Chain) const {
   const Function &F = DAG.getMachineFunction().getFunction();
 
   // First, check if tail calls have been disabled in this function.
-  if (F.getFnAttribute("disable-tail-calls").getValueAsBool())
+  if (F.getFnAttribute(DisableTailCallsAttr).getValueAsBool())
     return false;
 
   // Conservatively require the attributes of the call to match those of
   // the return. Ignore following attributes because they don't affect the
   // call sequence.
   AttrBuilder CallerAttrs(F.getAttributes(), AttributeList::ReturnIndex);
   for (const auto &Attr : {Attribute::Alignment, Attribute::Dereferenceable,
                            Attribute::DereferenceableOrNull, Attribute::NoAlias,

llvm/lib/CodeGen/StackProtector.cpp

DominatorTreeWrapperPass *DTWP =
       getAnalysisIfAvailable<DominatorTreeWrapperPass>();
   DT = DTWP ? &DTWP->getDomTree() : nullptr;
   TM = &getAnalysis<TargetPassConfig>().getTM<TargetMachine>();
   Trip = TM->getTargetTriple();
   TLI = TM->getSubtargetImpl(Fn)->getTargetLowering();
   HasPrologue = false;
   HasIRCheck = false;
 
-  Attribute Attr = Fn.getFnAttribute("stack-protector-buffer-size");
+  Attribute Attr = Fn.getFnAttribute(StackProtectorBufferSizeAttr);
   if (Attr.isStringAttribute() &&
       Attr.getValueAsString().getAsInteger(10, SSPBufferSize))
     return false; // Invalid integer string
 
   if (!RequiresStackProtector())
     return false;
 
   // TODO(etienneb): Functions with funclets are not correctly supported now.

llvm/lib/CodeGen/TargetLoweringBase.cpp

 //===----------------------------------------------------------------------===//
 //  Reciprocal Estimates
 //===----------------------------------------------------------------------===//
 
 /// Get the reciprocal estimate attribute string for a function that will
 /// override the target defaults.
 static StringRef getRecipEstimateForFunc(MachineFunction &MF) {
   const Function &F = MF.getFunction();
-  return F.getFnAttribute("reciprocal-estimates").getValueAsString();
+  return F.getFnAttribute(ReciprocalEstimatesAttr).getValueAsString();
 }
 
 /// Construct a string for the given reciprocal operation of the given type.
 /// This string should match the corresponding option to the front-end's
 /// "-mrecip" flag assuming those strings have been passed through in an
 /// attribute string. For example, "vec-divf" for a division of a vXf32.
 static std::string getReciprocalOpName(bool IsSqrt, EVT VT) {
   std::string Name = VT.isVector() ? "vec-" : "";

llvm/lib/CodeGen/TargetLoweringObjectFileImpl.cpp

static MCSection *selectExplicitSectionGlobal(
   StringRef SectionName = GO->getSection();
 
   // Check if '#pragma clang section' name is applicable.
   // Note that pragma directive overrides -ffunction-section, -fdata-section
   // and so section name is exactly as user specified and not uniqued.
   const GlobalVariable *GV = dyn_cast<GlobalVariable>(GO);
   if (GV && GV->hasImplicitSection()) {
     auto Attrs = GV->getAttributes();
-    if (Attrs.hasAttribute("bss-section") && Kind.isBSS()) {
-      SectionName = Attrs.getAttribute("bss-section").getValueAsString();
-    } else if (Attrs.hasAttribute("rodata-section") && Kind.isReadOnly()) {
-      SectionName = Attrs.getAttribute("rodata-section").getValueAsString();
-    } else if (Attrs.hasAttribute("relro-section") && Kind.isReadOnlyWithRel()) {
-      SectionName = Attrs.getAttribute("relro-section").getValueAsString();
-    } else if (Attrs.hasAttribute("data-section") && Kind.isData()) {
-      SectionName = Attrs.getAttribute("data-section").getValueAsString();
+    if (Attrs.hasAttribute(BssSectionAttr) && Kind.isBSS()) {
+      SectionName = Attrs.getAttribute(BssSectionAttr).getValueAsString();
+    } else if (Attrs.hasAttribute(RodataSectionAttr) && Kind.isReadOnly()) {
+      SectionName = Attrs.getAttribute(RodataSectionAttr).getValueAsString();
+    } else if (Attrs.hasAttribute(RelroSectionAttr) &&
+               Kind.isReadOnlyWithRel()) {
+      SectionName = Attrs.getAttribute(RelroSectionAttr).getValueAsString();
+    } else if (Attrs.hasAttribute(DataSectionAttr) && Kind.isData()) {
+      SectionName = Attrs.getAttribute(DataSectionAttr).getValueAsString();
     }
   }
   const Function *F = dyn_cast<Function>(GO);
-  if (F && F->hasFnAttribute("implicit-section-name")) {
-    SectionName = F->getFnAttribute("implicit-section-name").getValueAsString();
+  if (F && F->hasFnAttribute(ImplicitSectionNameAttr)) {
+    SectionName = F->getFnAttribute(ImplicitSectionNameAttr).getValueAsString();
   }
 
   // Infer section flags from the section name if we can.
   Kind = getELFKindForNamedSection(SectionName, Kind);
 
   StringRef Group = "";
   bool IsComdat = false;
   unsigned Flags = getELFSectionFlags(Kind);

 }
 
 MCSection *TargetLoweringObjectFileELF::getUniqueSectionForFunction(
     const Function &F, const TargetMachine &TM) const {
   SectionKind Kind = SectionKind::getText();
   unsigned Flags = getELFSectionFlags(Kind);
   // If the function's section names is pre-determined via pragma or a
   // section attribute, call selectExplicitSectionGlobal.
-  if (F.hasSection() || F.hasFnAttribute("implicit-section-name"))
+  if (F.hasSection() || F.hasFnAttribute(ImplicitSectionNameAttr))
     return selectExplicitSectionGlobal(
         &F, Kind, TM, getContext(), getMangler(), NextUniqueID,
         Used.count(&F), /* ForceUnique = */true);
   else
     return selectELFSectionForGlobal(
         getContext(), &F, Kind, getMangler(), TM, Used.count(&F),
         /*EmitUniqueSection=*/true, Flags, &NextUniqueID);
 }

 }
 
 MCSection *TargetLoweringObjectFileMachO::getExplicitSectionGlobal(
     const GlobalObject *GO, SectionKind Kind, const TargetMachine &TM) const {
 
   StringRef SectionName = GO->getSection();
 
   const Function *F = dyn_cast<Function>(GO);
-  if (F && F->hasFnAttribute("implicit-section-name")) {
-    SectionName = F->getFnAttribute("implicit-section-name").getValueAsString();
+  if (F && F->hasFnAttribute(ImplicitSectionNameAttr)) {
+    SectionName = F->getFnAttribute(ImplicitSectionNameAttr).getValueAsString();
   }
 
   // Parse the section specifier and create it if valid.
   StringRef Segment, Section;
   unsigned TAA = 0, StubSize = 0;
   bool TAAParsed;
 
   checkMachOComdat(GO);
TargetLoweringObjectFileXCOFF::getTargetSymbol(const GlobalValue *GV,
   // If a GV represents a GlobalVariable and -fdata-sections is enabled, we
   // also return a qualname so that a label symbol could be avoided.
   // It is inherently ambiguous when the GO represents the address of a
   // function, as the GO could either represent a function descriptor or a
   // function entry point. We choose to always return a function descriptor
   // here.
   if (const GlobalObject *GO = dyn_cast<GlobalObject>(GV)) {
     if (const GlobalVariable *GVar = dyn_cast<GlobalVariable>(GV))
-      if (GVar->hasAttribute("toc-data"))
+      if (GVar->hasAttribute(TocDataAttr))
         return cast<MCSectionXCOFF>(
                    SectionForGlobal(GVar, SectionKind::getData(), TM))
             ->getQualNameSymbol();
 
     if (GO->isDeclarationForLinker())
       return cast<MCSectionXCOFF>(getSectionForExternalReference(GO, TM))
           ->getQualNameSymbol();
 
MCSection *TargetLoweringObjectFileXCOFF::getExplicitSectionGlobal(
     const GlobalObject *GO, SectionKind Kind, const TargetMachine &TM) const {
   if (!GO->hasSection())
     report_fatal_error("#pragma clang section is not yet supported");
 
   StringRef SectionName = GO->getSection();
 
   // Handle the XCOFF::TD case first, then deal with the rest.
   if (const GlobalVariable *GVar = dyn_cast<GlobalVariable>(GO))
-    if (GVar->hasAttribute("toc-data"))
+    if (GVar->hasAttribute(TocDataAttr))
       return getContext().getXCOFFSection(
           SectionName, Kind,
           XCOFF::CsectProperties(/*MappingClass*/ XCOFF::XMC_TD, XCOFF::XTY_SD),
           /* MultiSymbolsAllowed*/ true);
 
   XCOFF::StorageMappingClass MappingClass;
   if (Kind.isText())
     MappingClass = XCOFF::XMC_PR;
return getContext().getXCOFFSection(
       Name, SectionKind::getMetadata(),
       XCOFF::CsectProperties(SMC, XCOFF::XTY_ER));
 }
 
 MCSection *TargetLoweringObjectFileXCOFF::SelectSectionForGlobal(
     const GlobalObject *GO, SectionKind Kind, const TargetMachine &TM) const {
   // Handle the XCOFF::TD case first, then deal with the rest.
   if (const GlobalVariable *GVar = dyn_cast<GlobalVariable>(GO))
-    if (GVar->hasAttribute("toc-data")) {
+    if (GVar->hasAttribute(TocDataAttr)) {
       SmallString<128> Name;
       getNameWithPrefix(Name, GO, TM);
       return getContext().getXCOFFSection(
           Name, Kind, XCOFF::CsectProperties(XCOFF::XMC_TD, XCOFF::XTY_SD),
           /* MultiSymbolsAllowed*/ true);
     }
 
   // Common symbols go into a csect with matching name which will get mapped

llvm/lib/CodeGen/TargetOptionsImpl.cpp

 /// optimization should be disabled for the given machine function.
 bool TargetOptions::DisableFramePointerElim(const MachineFunction &MF) const {
   // Check to see if the target want to forcably keep frame pointer.
   if (MF.getSubtarget().getFrameLowering()->keepFramePointer(MF))
     return true;
 
   const Function &F = MF.getFunction();
 
-  if (!F.hasFnAttribute("frame-pointer"))
+  if (!F.hasFnAttribute(FramePointerAttr))
     return false;
-  StringRef FP = F.getFnAttribute("frame-pointer").getValueAsString();
+  StringRef FP = F.getFnAttribute(FramePointerAttr).getValueAsString();
   if (FP == "all")
     return true;
   if (FP == "non-leaf")
     return MF.getFrameInfo().hasCalls();
   if (FP == "none")
     return false;
   llvm_unreachable("unknown frame pointer flag");
 }

llvm/lib/CodeGen/TargetRegisterInfo.cpp

if (callerPreservedRegs) {
     assert(Register::isPhysicalRegister(PhysReg) &&
            "Expected physical register");
     return (callerPreservedRegs[PhysReg / 32] >> PhysReg % 32) & 1;
   }
   return false;
 }
 
 bool TargetRegisterInfo::canRealignStack(const MachineFunction &MF) const {
-  return !MF.getFunction().hasFnAttribute("no-realign-stack");
+  return !MF.getFunction().hasFnAttribute(NoRealignStack);
 }
 
 bool TargetRegisterInfo::shouldRealignStack(const MachineFunction &MF) const {
   const MachineFrameInfo &MFI = MF.getFrameInfo();
   const TargetFrameLowering *TFI = MF.getSubtarget().getFrameLowering();
   const Function &F = MF.getFunction();
-  return F.hasFnAttribute("stackrealign") ||
+  return F.hasFnAttribute(StackrealignAttr) ||
          (MFI.getMaxAlign() > TFI->getStackAlign()) ||
          F.hasFnAttribute(Attribute::StackAlignment);
 }
 
 bool TargetRegisterInfo::regmaskSubsetEqual(const uint32_t *mask0,
                                             const uint32_t *mask1) const {
   unsigned N = (getNumRegs()+31) / 32;
   for (unsigned I = 0; I < N; ++I)

llvm/lib/CodeGen/XRayInstrumentation.cpp

for (auto &T : MBB.terminators()) {
         //   PATCHABLE_TAIL_CALL .
         BuildMI(MBB, T, T.getDebugLoc(), TII->get(Opc));
       }
     }
 }
 
 bool XRayInstrumentation::runOnMachineFunction(MachineFunction &MF) {
   auto &F = MF.getFunction();
-  auto InstrAttr = F.getFnAttribute("function-instrument");
+  auto InstrAttr = F.getFnAttribute(FunctionInstrumentAttr);
   bool AlwaysInstrument = InstrAttr.isStringAttribute() &&
                           InstrAttr.getValueAsString() == "xray-always";
   bool NeverInstrument = InstrAttr.isStringAttribute() &&
                          InstrAttr.getValueAsString() == "xray-never";
   if (NeverInstrument && !AlwaysInstrument)
     return false;
-  auto ThresholdAttr = F.getFnAttribute("xray-instruction-threshold");
-  auto IgnoreLoopsAttr = F.getFnAttribute("xray-ignore-loops");
+  auto ThresholdAttr = F.getFnAttribute(XrayInstructionThresholdAttr);
+  auto IgnoreLoopsAttr = F.getFnAttribute(XrayIgnoreLoopsAttr);
   unsigned int XRayThreshold = 0;
   if (!AlwaysInstrument) {
     if (!ThresholdAttr.isStringAttribute())
       return false; // XRay threshold attribute not found.
     if (ThresholdAttr.getValueAsString().getAsInteger(10, XRayThreshold))
       return false; // Invalid value for threshold.
 
     bool IgnoreLoops = IgnoreLoopsAttr.isValid();
bool XRayInstrumentation::runOnMachineFunction(MachineFunction &MF) {
   auto &FirstMI = *FirstMBB.begin();
 
   if (!MF.getSubtarget().isXRaySupported()) {
     FirstMI.emitError("An attempt to perform XRay instrumentation for an"
                       " unsupported target.");
     return false;
   }
 
-  if (!F.hasFnAttribute("xray-skip-entry")) {
+  if (!F.hasFnAttribute(XraySkipEntryAttr)) {
     // First, insert an PATCHABLE_FUNCTION_ENTER as the first instruction of the
     // MachineFunction.
     BuildMI(FirstMBB, FirstMI, FirstMI.getDebugLoc(),
             TII->get(TargetOpcode::PATCHABLE_FUNCTION_ENTER));
   }
 
-  if (!F.hasFnAttribute("xray-skip-exit")) {
+  if (!F.hasFnAttribute(XraySkipExitAttr)) {
     switch (MF.getTarget().getTargetTriple().getArch()) {
     case Triple::ArchType::arm:
     case Triple::ArchType::thumb:
     case Triple::ArchType::aarch64:
     case Triple::ArchType::mips:
     case Triple::ArchType::mipsel:
     case Triple::ArchType::mips64:
     case Triple::ArchType::mips64el: {

llvm/lib/ExecutionEngine/ExecutionEngineBindings.cpp

LLVMBool LLVMCreateMCJITCompilerForModule(
   std::unique_ptr<Module> Mod(unwrap(M));
 
   if (Mod)
     // Set function attribute "frame-pointer" based on
     // NoFramePointerElim.
     for (auto &F : *Mod) {
       auto Attrs = F.getAttributes();
       StringRef Value = options.NoFramePointerElim ? "all" : "none";
-      Attrs = Attrs.addFnAttribute(F.getContext(), "frame-pointer", Value);
+      Attrs = Attrs.addFnAttribute(F.getContext(), FramePointerAttr, Value);
       F.setAttributes(Attrs);
     }
 
   std::string Error;
   EngineBuilder builder(std::move(Mod));
   builder.setEngineKind(EngineKind::JIT)
          .setErrorStr(&Error)
          .setOptLevel((CodeGenOpt::Level)options.OptLevel)

llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp

 /// or at least cache it between methods. Note that while fontends such as Clang
 /// have just a single main TargetMachine per translation unit, "target-cpu" and
 /// "target-features" that determine the TargetMachine are per-function and can
 /// be overrided using __attribute__((target("OPTIONS"))).
 static std::unique_ptr<TargetMachine>
 createTargetMachine(Function *F, CodeGenOpt::Level OptLevel) {
   Module *M = F->getParent();
 
-  StringRef CPU = F->getFnAttribute("target-cpu").getValueAsString();
-  StringRef Features = F->getFnAttribute("target-features").getValueAsString();
+  StringRef CPU = F->getFnAttribute(TargetCPUAttr).getValueAsString();
+  StringRef Features = F->getFnAttribute(TargetFeaturesAttr).getValueAsString();
   const std::string &Triple = M->getTargetTriple();
 
   std::string Error;
   const llvm::Target *TheTarget = TargetRegistry::lookupTarget(Triple, Error);
   if (!TheTarget)
     return {};
 
   llvm::TargetOptions Options;

llvm/lib/IR/Assumptions.cpp

bool addAssumptionsImpl(AttrSite &Site,
 
   DenseSet<StringRef> CurAssumptions = getAssumptions(Site);
 
   if (!set_union(CurAssumptions, Assumptions))
     return false;
 
   LLVMContext &Ctx = Site.getContext();
   Site.addFnAttr(llvm::Attribute::get(
-      Ctx, llvm::AssumptionAttrKey,
+      Ctx, llvm::LLVMAssumeAttr,
       llvm::join(CurAssumptions.begin(), CurAssumptions.end(), ",")));
 
   return true;
 }
 } // namespace
 
 bool llvm::hasAssumption(const Function &F,
                          const KnownAssumptionString &AssumptionStr) {
-  const Attribute &A = F.getFnAttribute(AssumptionAttrKey);
+  const Attribute &A = F.getFnAttribute(LLVMAssumeAttr);
   return ::hasAssumption(A, AssumptionStr);
 }
 
 bool llvm::hasAssumption(const CallBase &CB,
                          const KnownAssumptionString &AssumptionStr) {
   if (Function *F = CB.getCalledFunction())
     if (hasAssumption(*F, AssumptionStr))
       return true;
 
-  const Attribute &A = CB.getFnAttr(AssumptionAttrKey);
+  const Attribute &A = CB.getFnAttr(LLVMAssumeAttr);
   return ::hasAssumption(A, AssumptionStr);
 }
 
 DenseSet<StringRef> llvm::getAssumptions(const Function &F) {
-  const Attribute &A = F.getFnAttribute(AssumptionAttrKey);
+  const Attribute &A = F.getFnAttribute(LLVMAssumeAttr);
   return ::getAssumptions(A);
 }
 
 DenseSet<StringRef> llvm::getAssumptions(const CallBase &CB) {
-  const Attribute &A = CB.getFnAttr(AssumptionAttrKey);
+  const Attribute &A = CB.getFnAttr(LLVMAssumeAttr);
   return ::getAssumptions(A);
 }
 
 bool llvm::addAssumptions(Function &F, const DenseSet<StringRef> &Assumptions) {
   return ::addAssumptionsImpl(F, Assumptions);
 }
 
 bool llvm::addAssumptions(CallBase &CB,

llvm/lib/IR/AttributeImpl.h

public:
   AttributeImpl &operator=(const AttributeImpl &) = delete;
 
   bool isEnumAttribute() const { return KindID == EnumAttrEntry; }
   bool isIntAttribute() const { return KindID == IntAttrEntry; }
   bool isStringAttribute() const { return KindID == StringAttrEntry; }
   bool isTypeAttribute() const { return KindID == TypeAttrEntry; }
 
   bool hasAttribute(Attribute::AttrKind A) const;
-  bool hasAttribute(StringRef Kind) const;
+  bool hasAttribute(AttributeKey Kind) const;
 
   Attribute::AttrKind getKindAsEnum() const;
   uint64_t getValueAsInt() const;
   bool getValueAsBool() const;
 
-  StringRef getKindAsString() const;
+  AttributeKey getKindAsKey() const;
   StringRef getValueAsString() const;
 
   Type *getValueAsType() const;
 
   /// Used when sorting the attributes.
   bool operator<(const AttributeImpl &AI) const;
 
   void Profile(FoldingSetNodeID &ID) const {
     if (isEnumAttribute())
       Profile(ID, getKindAsEnum(), static_cast<uint64_t>(0));
     else if (isIntAttribute())
       Profile(ID, getKindAsEnum(), getValueAsInt());
     else if (isStringAttribute())
-      Profile(ID, getKindAsString(), getValueAsString());
+      Profile(ID, getKindAsKey(), getValueAsString());
     else
       Profile(ID, getKindAsEnum(), getValueAsType());
   }
 
   static void Profile(FoldingSetNodeID &ID, Attribute::AttrKind Kind,
                       uint64_t Val) {
     ID.AddInteger(Kind);
     if (Val) ID.AddInteger(Val);
   }
 
-  static void Profile(FoldingSetNodeID &ID, StringRef Kind, StringRef Values) {
-    ID.AddString(Kind);
+  static void Profile(FoldingSetNodeID &ID, AttributeKey Kind,
+                      StringRef Values) {
+    ID.AddString(Kind.value());
     if (!Values.empty()) ID.AddString(Values);
   }
 
   static void Profile(FoldingSetNodeID &ID, Attribute::AttrKind Kind,
                       Type *Ty) {
     ID.AddInteger(Kind);
     ID.AddPointer(Ty);
   }
public:
   uint64_t getValue() const { return Val; }
 };
 
 class StringAttributeImpl final
     : public AttributeImpl,
       private TrailingObjects<StringAttributeImpl, char> {
   friend TrailingObjects;
 
-  unsigned KindSize;
+  AttributeKey Kind;
   unsigned ValSize;
-  size_t numTrailingObjects(OverloadToken<char>) const {
-    return KindSize + 1 + ValSize + 1;
-  }
+  size_t numTrailingObjects(OverloadToken<char>) const { return ValSize + 1; }
 
 public:
-  StringAttributeImpl(StringRef Kind, StringRef Val = StringRef())
-      : AttributeImpl(StringAttrEntry), KindSize(Kind.size()),
-        ValSize(Val.size()) {
+  StringAttributeImpl(AttributeKey Kind, StringRef Val = StringRef())
+      : AttributeImpl(StringAttrEntry), Kind(Kind), ValSize(Val.size()) {
     char *TrailingString = getTrailingObjects<char>();
     // Some users rely on zero-termination.
-    llvm::copy(Kind, TrailingString);
-    TrailingString[KindSize] = '\0';
-    llvm::copy(Val, &TrailingString[KindSize + 1]);
-    TrailingString[KindSize + 1 + ValSize] = '\0';
+    llvm::copy(Val, TrailingString);
+    TrailingString[ValSize] = '\0';
   }
 
-  StringRef getStringKind() const {
-    return StringRef(getTrailingObjects<char>(), KindSize);
-  }
+  AttributeKey getStringKind() const { return Kind; }
   StringRef getStringValue() const {
-    return StringRef(getTrailingObjects<char>() + KindSize + 1, ValSize);
+    return StringRef(getTrailingObjects<char>(), ValSize);
   }
 
-  static size_t totalSizeToAlloc(StringRef Kind, StringRef Val) {
-    return TrailingObjects::totalSizeToAlloc<char>(Kind.size() + 1 +
-                                                   Val.size() + 1);
+  static size_t totalSizeToAlloc(StringRef Val) {
+    return TrailingObjects::totalSizeToAlloc<char>(Val.size() + 1);
   }
 };
 
 class TypeAttributeImpl : public EnumAttributeImpl {
   Type *Ty;
 
 public:
   TypeAttributeImpl(Attribute::AttrKind Kind, Type *Ty)

 class AttributeSetNode final
     : public FoldingSetNode,
       private TrailingObjects<AttributeSetNode, Attribute> {
   friend TrailingObjects;
 
   unsigned NumAttrs; ///< Number of attributes in this node.
   AttributeBitSet AvailableAttrs; ///< Available enum attributes.
 
-  DenseMap<StringRef, Attribute> StringAttrs;
+  std::unordered_map<AttributeKey, Attribute> StringAttrs;
 
   AttributeSetNode(ArrayRef<Attribute> Attrs);
 
   static AttributeSetNode *getSorted(LLVMContext &C,
                                      ArrayRef<Attribute> SortedAttrs);
   Optional<Attribute> findEnumAttribute(Attribute::AttrKind Kind) const;
 
 public:
   // AttributesSetNode is uniqued, these should not be available.
   AttributeSetNode(const AttributeSetNode &) = delete;
   AttributeSetNode &operator=(const AttributeSetNode &) = delete;
 
   void operator delete(void *p) { ::operator delete(p); }
 
   static AttributeSetNode *get(LLVMContext &C, const AttrBuilder &B);
 
   static AttributeSetNode *get(LLVMContext &C, ArrayRef<Attribute> Attrs);
 
   /// Return the number of attributes this AttributeList contains.
   unsigned getNumAttributes() const { return NumAttrs; }
 
   bool hasAttribute(Attribute::AttrKind Kind) const {
     return AvailableAttrs.hasAttribute(Kind);
   }
-  bool hasAttribute(StringRef Kind) const;
+  bool hasAttribute(AttributeKey Kind) const;
   bool hasAttributes() const { return NumAttrs != 0; }
 
   Attribute getAttribute(Attribute::AttrKind Kind) const;
-  Attribute getAttribute(StringRef Kind) const;
+  Attribute getAttribute(AttributeKey Kind) const;
 
   MaybeAlign getAlignment() const;
   MaybeAlign getStackAlignment() const;
   uint64_t getDereferenceableBytes() const;
   uint64_t getDereferenceableOrNullBytes() const;
   std::pair<unsigned, Optional<unsigned>> getAllocSizeArgs() const;
   std::pair<unsigned, unsigned> getVScaleRangeArgs() const;
   std::string getAsString(bool InAttrGrp) const;

llvm/lib/IR/Attributes.cpp

else
       PA = new (pImpl->Alloc) IntAttributeImpl(Kind, Val);
     pImpl->AttrsSet.InsertNode(PA, InsertPoint);
   }
 
   // Return the Attribute that we found or created.
   return Attribute(PA);
 }
 
-Attribute Attribute::get(LLVMContext &Context, StringRef Kind, StringRef Val) {
+Attribute Attribute::get(LLVMContext &Context, AttributeKey Kind,
+                         StringRef Val) {
   LLVMContextImpl *pImpl = Context.pImpl;
   FoldingSetNodeID ID;
-  ID.AddString(Kind);
+  ID.AddString(Kind.value());
   if (!Val.empty()) ID.AddString(Val);
 
   void *InsertPoint;
   AttributeImpl *PA = pImpl->AttrsSet.FindNodeOrInsertPos(ID, InsertPoint);
 
   if (!PA) {
     // If we didn't find any existing attributes of the same shape then create a
     // new one and insert it.
     void *Mem =
-        pImpl->Alloc.Allocate(StringAttributeImpl::totalSizeToAlloc(Kind, Val),
+        pImpl->Alloc.Allocate(StringAttributeImpl::totalSizeToAlloc(Val),
                               alignof(StringAttributeImpl));
     PA = new (Mem) StringAttributeImpl(Kind, Val);
     pImpl->AttrsSet.InsertNode(PA, InsertPoint);
   }
 
   // Return the Attribute that we found or created.
   return Attribute(PA);
 }

 
 bool Attribute::getValueAsBool() const {
   if (!pImpl) return false;
   assert(isStringAttribute() &&
          "Expected the attribute to be a string attribute!");
   return pImpl->getValueAsBool();
 }
 
-StringRef Attribute::getKindAsString() const {
+AttributeKey Attribute::getKindAsKey() const {
   if (!pImpl) return {};
   assert(isStringAttribute() &&
          "Invalid attribute type to get the kind as a string!");
-  return pImpl->getKindAsString();
+  return pImpl->getKindAsKey();
 }
 
 StringRef Attribute::getValueAsString() const {
   if (!pImpl) return {};
   assert(isStringAttribute() &&
          "Invalid attribute type to get the value as a string!");
   return pImpl->getValueAsString();
 }
 
 Type *Attribute::getValueAsType() const {
   if (!pImpl) return {};
   assert(isTypeAttribute() &&
          "Invalid attribute type to get the value as a type!");
   return pImpl->getValueAsType();
 }
 
 
 bool Attribute::hasAttribute(AttrKind Kind) const {
   return (pImpl && pImpl->hasAttribute(Kind)) || (!pImpl && Kind == None);
 }
 
-bool Attribute::hasAttribute(StringRef Kind) const {
+bool Attribute::hasAttribute(AttributeKey Kind) const {
   if (!isStringAttribute()) return false;
   return pImpl && pImpl->hasAttribute(Kind);
 }
 
 MaybeAlign Attribute::getAlignment() const {
   assert(hasAttribute(Attribute::Alignment) &&
          "Trying to get alignment from non-alignment attribute!");
   return MaybeAlign(pImpl->getValueAsInt());
std::string Attribute::getAsString(bool InAttrGrp) const {
   //
   //   "kind"
   //   "kind" = "value"
   //
   if (isStringAttribute()) {
     std::string Result;
     {
       raw_string_ostream OS(Result);
-      OS << '"' << getKindAsString() << '"';
+      OS << '"' << getKindAsKey().value() << '"';
 
       // Since some attribute strings contain special characters that cannot be
       // printable, those have to be escaped to make the attribute value
       // printable as is.  e.g. "\01__gnu_mcount_nc"
       const auto &AttrVal = pImpl->getValueAsString();
       if (!AttrVal.empty()) {
         OS << "=\"";
         printEscapedString(AttrVal, OS);

 // AttributeImpl Definition
 //===----------------------------------------------------------------------===//
 
 bool AttributeImpl::hasAttribute(Attribute::AttrKind A) const {
   if (isStringAttribute()) return false;
   return getKindAsEnum() == A;
 }
 
-bool AttributeImpl::hasAttribute(StringRef Kind) const {
+bool AttributeImpl::hasAttribute(AttributeKey Kind) const {
   if (!isStringAttribute()) return false;
-  return getKindAsString() == Kind;
+  return getKindAsKey() == Kind;
 }
 
 Attribute::AttrKind AttributeImpl::getKindAsEnum() const {
   assert(isEnumAttribute() || isIntAttribute() || isTypeAttribute());
   return static_cast<const EnumAttributeImpl *>(this)->getEnumKind();
 }
 
 uint64_t AttributeImpl::getValueAsInt() const {
   assert(isIntAttribute());
   return static_cast<const IntAttributeImpl *>(this)->getValue();
 }
 
 bool AttributeImpl::getValueAsBool() const {
   assert(getValueAsString().empty() || getValueAsString() == "false" || getValueAsString() == "true");
   return getValueAsString() == "true";
 }
 
-StringRef AttributeImpl::getKindAsString() const {
+AttributeKey AttributeImpl::getKindAsKey() const {
   assert(isStringAttribute());
   return static_cast<const StringAttributeImpl *>(this)->getStringKind();
 }
 
 StringRef AttributeImpl::getValueAsString() const {
   assert(isStringAttribute());
   return static_cast<const StringAttributeImpl *>(this)->getStringValue();
 }
if (!isStringAttribute()) {
     assert(!AI.isTypeAttribute() && "Comparison of types would be unstable");
     // TODO: Is this actually needed?
     assert(AI.isIntAttribute() && "Only possibility left");
     return getValueAsInt() < AI.getValueAsInt();
   }
 
   if (!AI.isStringAttribute())
     return false;
-  if (getKindAsString() == AI.getKindAsString())
+  if (getKindAsKey() == AI.getKindAsKey())
     return getValueAsString() < AI.getValueAsString();
-  return getKindAsString() < AI.getKindAsString();
+  return getKindAsKey() < AI.getKindAsKey();
 }
 
 //===----------------------------------------------------------------------===//
 // AttributeSet Definition
 //===----------------------------------------------------------------------===//
 
 AttributeSet AttributeSet::get(LLVMContext &C, const AttrBuilder &B) {
   return AttributeSet(AttributeSetNode::get(C, B));
 }
 
 AttributeSet AttributeSet::get(LLVMContext &C, ArrayRef<Attribute> Attrs) {
   return AttributeSet(AttributeSetNode::get(C, Attrs));
 }
 
 AttributeSet AttributeSet::addAttribute(LLVMContext &C,
                                         Attribute::AttrKind Kind) const {
   if (hasAttribute(Kind)) return *this;
   AttrBuilder B;
   B.addAttribute(Kind);
   return addAttributes(C, AttributeSet::get(C, B));
 }
 
-AttributeSet AttributeSet::addAttribute(LLVMContext &C, StringRef Kind,
+AttributeSet AttributeSet::addAttribute(LLVMContext &C, AttributeKey Kind,
                                         StringRef Value) const {
   AttrBuilder B;
   B.addAttribute(Kind, Value);
   return addAttributes(C, AttributeSet::get(C, B));
 }
 
 AttributeSet AttributeSet::addAttributes(LLVMContext &C,
                                          const AttributeSet AS) const {
AttributeSet AttributeSet::removeAttribute(LLVMContext &C,
                                              Attribute::AttrKind Kind) const {
   if (!hasAttribute(Kind)) return *this;
   AttrBuilder B(*this);
   B.removeAttribute(Kind);
   return get(C, B);
 }
 
 AttributeSet AttributeSet::removeAttribute(LLVMContext &C,
-                                             StringRef Kind) const {
+                                           AttributeKey Kind) const {
   if (!hasAttribute(Kind)) return *this;
   AttrBuilder B(*this);
   B.removeAttribute(Kind);
   return get(C, B);
 }
 
 AttributeSet AttributeSet::removeAttributes(LLVMContext &C,
                                             const AttrBuilder &Attrs) const {

 unsigned AttributeSet::getNumAttributes() const {
   return SetNode ? SetNode->getNumAttributes() : 0;
 }
 
 bool AttributeSet::hasAttribute(Attribute::AttrKind Kind) const {
   return SetNode ? SetNode->hasAttribute(Kind) : false;
 }
 
-bool AttributeSet::hasAttribute(StringRef Kind) const {
+bool AttributeSet::hasAttribute(AttributeKey Kind) const {
   return SetNode ? SetNode->hasAttribute(Kind) : false;
 }
 
 Attribute AttributeSet::getAttribute(Attribute::AttrKind Kind) const {
   return SetNode ? SetNode->getAttribute(Kind) : Attribute();
 }
 
-Attribute AttributeSet::getAttribute(StringRef Kind) const {
+Attribute AttributeSet::getAttribute(AttributeKey Kind) const {
   return SetNode ? SetNode->getAttribute(Kind) : Attribute();
 }
 
 MaybeAlign AttributeSet::getAlignment() const {
   return SetNode ? SetNode->getAlignment() : None;
 }
 
 MaybeAlign AttributeSet::getStackAlignment() const {

 
 AttributeSetNode::AttributeSetNode(ArrayRef<Attribute> Attrs)
     : NumAttrs(Attrs.size()) {
   // There's memory after the node where we can store the entries in.
   llvm::copy(Attrs, getTrailingObjects<Attribute>());
 
   for (const auto &I : *this) {
     if (I.isStringAttribute())
-      StringAttrs.insert({ I.getKindAsString(), I });
+      StringAttrs.insert({I.getKindAsKey(), I});
     else
       AvailableAttrs.addAttribute(I.getKindAsEnum());
   }
 }
 
 AttributeSetNode *AttributeSetNode::get(LLVMContext &C,
                                         ArrayRef<Attribute> Attrs) {
   SmallVector<Attribute, 8> SortedAttrs(Attrs.begin(), Attrs.end());
AttributeSetNode *AttributeSetNode::get(LLVMContext &C, const AttrBuilder &B) {
 
   // Add target-dependent (string) attributes.
   for (const auto &TDA : B.td_attrs())
     Attrs.emplace_back(Attribute::get(C, TDA.first, TDA.second));
 
   return getSorted(C, Attrs);
 }
 
-bool AttributeSetNode::hasAttribute(StringRef Kind) const {
+bool AttributeSetNode::hasAttribute(AttributeKey Kind) const {
   return StringAttrs.count(Kind);
 }
 
 Optional<Attribute>
 AttributeSetNode::findEnumAttribute(Attribute::AttrKind Kind) const {
   // Do a quick presence check.
   if (!hasAttribute(Kind))
     return None;

 }
 
 Attribute AttributeSetNode::getAttribute(Attribute::AttrKind Kind) const {
   if (auto A = findEnumAttribute(Kind))
     return *A;
   return {};
 }
 
-Attribute AttributeSetNode::getAttribute(StringRef Kind) const {
-  return StringAttrs.lookup(Kind);
+Attribute AttributeSetNode::getAttribute(AttributeKey Kind) const {
+  auto Where = StringAttrs.find(Kind);
+  return Where != StringAttrs.end() ? Where->second : Attribute();
 }
 
 MaybeAlign AttributeSetNode::getAlignment() const {
   if (auto A = findEnumAttribute(Attribute::Alignment))
     return A->getAlignment();
   return None;
 }
 
AttributeList AttributeList::get(LLVMContext &C, unsigned Index,
   SmallVector<std::pair<unsigned, Attribute>, 8> Attrs;
   auto VI = Values.begin();
   for (const auto K : Kinds)
     Attrs.emplace_back(Index, Attribute::get(C, K, *VI++));
   return get(C, Attrs);
 }
 
 AttributeList AttributeList::get(LLVMContext &C, unsigned Index,
-                                 ArrayRef<StringRef> Kinds) {
+                                 ArrayRef<AttributeKey> Kinds) {
   SmallVector<std::pair<unsigned, Attribute>, 8> Attrs;
   for (const auto &K : Kinds)
     Attrs.emplace_back(Index, Attribute::get(C, K));
   return get(C, Attrs);
 }
 
 AttributeList AttributeList::get(LLVMContext &C,
                                  ArrayRef<AttributeList> Attrs) {
AttributeList::addAttributeAtIndex(LLVMContext &C, unsigned Index,
   AttributeSet Attrs = getAttributes(Index);
   // TODO: Insert at correct position and avoid sort.
   SmallVector<Attribute, 8> NewAttrs(Attrs.begin(), Attrs.end());
   NewAttrs.push_back(Attribute::get(C, Kind));
   return setAttributesAtIndex(C, Index, AttributeSet::get(C, NewAttrs));
 }
 
 AttributeList AttributeList::addAttributeAtIndex(LLVMContext &C, unsigned Index,
-                                                 StringRef Kind,
+                                                 AttributeKey Kind,
                                                  StringRef Value) const {
   AttrBuilder B;
   B.addAttribute(Kind, Value);
   return addAttributesAtIndex(C, Index, B);
 }
 
 AttributeList AttributeList::addAttributeAtIndex(LLVMContext &C, unsigned Index,
                                                  Attribute A) const {
AttributeList::removeAttributeAtIndex(LLVMContext &C, unsigned Index,
 
   AttrSets[Index] = AttrSets[Index].removeAttribute(C, Kind);
 
   return getImpl(C, AttrSets);
 }
 
 AttributeList AttributeList::removeAttributeAtIndex(LLVMContext &C,
                                                     unsigned Index,
-                                                    StringRef Kind) const {
+                                                    AttributeKey Kind) const {
   if (!hasAttributeAtIndex(Index, Kind))
     return *this;
 
   Index = attrIdxToArrayIdx(Index);
   SmallVector<AttributeSet, 4> AttrSets(this->begin(), this->end());
   assert(Index < AttrSets.size());
 
   AttrSets[Index] = AttrSets[Index].removeAttribute(C, Kind);
AttributeSet AttributeList::getFnAttrs() const {
   return getAttributes(FunctionIndex);
 }
 
 bool AttributeList::hasAttributeAtIndex(unsigned Index,
                                         Attribute::AttrKind Kind) const {
   return getAttributes(Index).hasAttribute(Kind);
 }
 
-bool AttributeList::hasAttributeAtIndex(unsigned Index, StringRef Kind) const {
+bool AttributeList::hasAttributeAtIndex(unsigned Index,
+                                        AttributeKey Kind) const {
   return getAttributes(Index).hasAttribute(Kind);
 }
 
 bool AttributeList::hasAttributesAtIndex(unsigned Index) const {
   return getAttributes(Index).hasAttributes();
 }
 
 bool AttributeList::hasFnAttr(Attribute::AttrKind Kind) const {
   return pImpl && pImpl->hasFnAttribute(Kind);
 }
 
-bool AttributeList::hasFnAttr(StringRef Kind) const {
+bool AttributeList::hasFnAttr(AttributeKey Kind) const {
   return hasAttributeAtIndex(AttributeList::FunctionIndex, Kind);
 }
 
 bool AttributeList::hasAttrSomewhere(Attribute::AttrKind Attr,
                                      unsigned *Index) const {
   return pImpl && pImpl->hasAttrSomewhere(Attr, Index);
 }
 
 Attribute AttributeList::getAttributeAtIndex(unsigned Index,
                                              Attribute::AttrKind Kind) const {
   return getAttributes(Index).getAttribute(Kind);
 }
 
 Attribute AttributeList::getAttributeAtIndex(unsigned Index,
-                                             StringRef Kind) const {
+                                             AttributeKey Kind) const {
   return getAttributes(Index).getAttribute(Kind);
 }
 
 MaybeAlign AttributeList::getRetAlignment() const {
   return getAttributes(ReturnIndex).getAlignment();
 }
 
 MaybeAlign AttributeList::getParamAlignment(unsigned ArgNo) const {

 AttrBuilder::kindToTypeIndex(Attribute::AttrKind Kind) const {
   if (Attribute::isTypeAttrKind(Kind))
     return Kind - Attribute::FirstTypeAttr;
   return None;
 }
 
 AttrBuilder &AttrBuilder::addAttribute(Attribute Attr) {
   if (Attr.isStringAttribute()) {
-    addAttribute(Attr.getKindAsString(), Attr.getValueAsString());
+    addAttribute(Attr.getKindAsKey(), Attr.getValueAsString());
     return *this;
   }
 
   Attribute::AttrKind Kind = Attr.getKindAsEnum();
   Attrs[Kind] = true;
 
   if (Optional<unsigned> TypeIndex = kindToTypeIndex(Kind))
     TypeAttrs[*TypeIndex] = Attr.getValueAsType();
   else if (Optional<unsigned> IntIndex = kindToIntIndex(Kind))
     IntAttrs[*IntIndex] = Attr.getValueAsInt();
 
   return *this;
 }
 
-AttrBuilder &AttrBuilder::addAttribute(StringRef A, StringRef V) {
+AttrBuilder &AttrBuilder::addAttribute(AttributeKey A, StringRef V) {
   TargetDepAttrs[A] = V;
   return *this;
 }
 
 AttrBuilder &AttrBuilder::removeAttribute(Attribute::AttrKind Val) {
   assert((unsigned)Val < Attribute::EndAttrKinds && "Attribute out of range!");
   Attrs[Val] = false;
 
   if (Optional<unsigned> TypeIndex = kindToTypeIndex(Val))
     TypeAttrs[*TypeIndex] = nullptr;
   else if (Optional<unsigned> IntIndex = kindToIntIndex(Val))
     IntAttrs[*IntIndex] = 0;
 
   return *this;
 }
 
 AttrBuilder &AttrBuilder::removeAttributes(AttributeList A, uint64_t Index) {
   remove(A.getAttributes(Index));
   return *this;
 }
 
-AttrBuilder &AttrBuilder::removeAttribute(StringRef A) {
+AttrBuilder &AttrBuilder::removeAttribute(AttributeKey A) {
   TargetDepAttrs.erase(A);
   return *this;
 }
 
 uint64_t AttrBuilder::getRawIntAttr(Attribute::AttrKind Kind) const {
   Optional<unsigned> IntIndex = kindToIntIndex(Kind);
   assert(IntIndex && "Not an int attribute");
   return IntAttrs[*IntIndex];
bool AttrBuilder::overlaps(const AttrBuilder &B) const {
   // Then check if any target dependent ones do.
   for (const auto &I : td_attrs())
     if (B.contains(I.first))
       return true;
 
   return false;
 }
 
-bool AttrBuilder::contains(StringRef A) const {
+bool AttrBuilder::contains(AttributeKey A) const {
   return TargetDepAttrs.find(A) != TargetDepAttrs.end();
 }
 
 bool AttrBuilder::hasAttributes() const {
   return !Attrs.none() || !TargetDepAttrs.empty();
 }
 
 bool AttrBuilder::hasAttributes(AttributeList AL, uint64_t Index) const {
   AttributeSet AS = AL.getAttributes(Index);
 
   for (const auto &Attr : AS) {
     if (Attr.isEnumAttribute() || Attr.isIntAttribute()) {
       if (contains(Attr.getKindAsEnum()))
         return true;
     } else {
       assert(Attr.isStringAttribute() && "Invalid attribute kind!");
-      return contains(Attr.getKindAsString());
+      return contains(Attr.getKindAsKey());
     }
   }
 
   return false;
 }
 
 bool AttrBuilder::hasAlignmentAttr() const {
   return getRawIntAttr(Attribute::Alignment) != 0;
Caller.addFnAttr(Attribute::StackProtectStrong);
              !Caller.hasFnAttribute(Attribute::StackProtectReq) &&
              !Caller.hasFnAttribute(Attribute::StackProtectStrong))
     Caller.addFnAttr(Attribute::StackProtect);
 }
 
 /// If the inlined function required stack probes, then ensure that
 /// the calling function has those too.
 static void adjustCallerStackProbes(Function &Caller, const Function &Callee) {
-  if (!Caller.hasFnAttribute("probe-stack") &&
-      Callee.hasFnAttribute("probe-stack")) {
-    Caller.addFnAttr(Callee.getFnAttribute("probe-stack"));
+  constexpr AttributeKey ProbeStackAttr("probe-stack");
+  if (!Caller.hasFnAttribute(ProbeStackAttr) &&
+      Callee.hasFnAttribute(ProbeStackAttr)) {
+    Caller.addFnAttr(Callee.getFnAttribute(ProbeStackAttr));
   }
 }
 
 /// If the inlined function defines the size of guard region
 /// on the stack, then ensure that the calling function defines a guard region
 /// that is no larger.
 static void
 adjustCallerStackProbeSize(Function &Caller, const Function &Callee) {
-  Attribute CalleeAttr = Callee.getFnAttribute("stack-probe-size");
+  constexpr AttributeKey StackProbeSizeAttr("stack-probe-size");
+  Attribute CalleeAttr = Callee.getFnAttribute(StackProbeSizeAttr);
   if (CalleeAttr.isValid()) {
-    Attribute CallerAttr = Caller.getFnAttribute("stack-probe-size");
+    Attribute CallerAttr = Caller.getFnAttribute(StackProbeSizeAttr);
     if (CallerAttr.isValid()) {
       uint64_t CallerStackProbeSize, CalleeStackProbeSize;
       CallerAttr.getValueAsString().getAsInteger(0, CallerStackProbeSize);
       CalleeAttr.getValueAsString().getAsInteger(0, CalleeStackProbeSize);
 
       if (CallerStackProbeSize > CalleeStackProbeSize) {
         Caller.addFnAttr(CalleeAttr);
       }

 /// attribute from the caller since we can't make any guarantees about the
 /// caller's requirements.
 /// This function is called after the inlining decision has been made so we have
 /// to merge the attribute this way. Heuristics that would use
 /// min-legal-vector-width to determine inline compatibility would need to be
 /// handled as part of inline cost analysis.
 static void
 adjustMinLegalVectorWidth(Function &Caller, const Function &Callee) {
-  Attribute CallerAttr = Caller.getFnAttribute("min-legal-vector-width");
+  constexpr AttributeKey AttrK("min-legal-vector-width");
+  Attribute CallerAttr = Caller.getFnAttribute(AttrK);
   if (CallerAttr.isValid()) {
-    Attribute CalleeAttr = Callee.getFnAttribute("min-legal-vector-width");
+    Attribute CalleeAttr = Callee.getFnAttribute(AttrK);
     if (CalleeAttr.isValid()) {
       uint64_t CallerVectorWidth, CalleeVectorWidth;
       CallerAttr.getValueAsString().getAsInteger(0, CallerVectorWidth);
       CalleeAttr.getValueAsString().getAsInteger(0, CalleeVectorWidth);
       if (CallerVectorWidth < CalleeVectorWidth)
         Caller.addFnAttr(CalleeAttr);
     } else {
       // If the callee doesn't have the attribute then we don't know anything
       // and must drop the attribute from the caller.
-      Caller.removeFnAttr("min-legal-vector-width");
+      Caller.removeFnAttr(AttrK);
     }
   }
 }
 
 /// If the inlined function has null_pointer_is_valid attribute,
 /// set this attribute in the caller post inlining.
 static void
 adjustNullPointerValidAttr(Function &Caller, const Function &Callee) {
static void set(Function &Fn,
     if (Val)
       Fn.addFnAttr(Kind);
     else
       Fn.removeFnAttr(Kind);
   }
 };
 
 struct StrBoolAttr {
-  static bool isSet(const Function &Fn,
-                    StringRef Kind) {
+  static bool isSet(const Function &Fn, AttributeKey Kind) {
     auto A = Fn.getFnAttribute(Kind);
     return A.getValueAsString().equals("true");
   }
 
-  static void set(Function &Fn,
-                  StringRef Kind, bool Val) {
+  static void set(Function &Fn, AttributeKey Kind, bool Val) {
     Fn.addFnAttr(Kind, Val ? "true" : "false");
   }
 };
 
 #define GET_ATTR_NAMES
 #define ATTRIBUTE_ENUM(ENUM_NAME, DISPLAY_NAME)                                \
-  struct ENUM_NAME##Attr : EnumAttr {                                          \
+  struct ENUM_NAME##_Attr : EnumAttr {                                         \
     static enum Attribute::AttrKind getKind() {                                \
       return llvm::Attribute::ENUM_NAME;                                       \
     }                                                                          \
   };
 #define ATTRIBUTE_STRBOOL(ENUM_NAME, DISPLAY_NAME)                             \
-  struct ENUM_NAME##Attr : StrBoolAttr {                                       \
-    static StringRef getKind() { return #DISPLAY_NAME; }                       \
+  struct ENUM_NAME##_Attr : StrBoolAttr {                                      \
+    static AttributeKey getKind() { return ENUM_NAME##Attr; }                  \
   };
 #include "llvm/IR/Attributes.inc"
 
 #define GET_ATTR_COMPAT_FUNC
 #include "llvm/IR/Attributes.inc"
 
 bool AttributeFuncs::areInlineCompatible(const Function &Caller,
                                          const Function &Callee) {
void AttributeFuncs::mergeAttributesForOutlining(Function &Base,
   // For example, if the NoNansFPMathAttr is set in one function, but not in
   // the other, in the merged function we can say that the NoNansFPMathAttr
   // is not set.
   // However if we have the SpeculativeLoadHardeningAttr set true in one
   // function, but not the other, we make sure that the function retains
   // that aspect in the merged function.
   mergeFnAttrs(Base, ToMerge);
 }
+
+static DenseMap<StringRef, AttributeKey> AttributeKeyPool;
+
+AttributeKey AttributeKey::Create(StringRef s) {
+  assert(s.size() < std::numeric_limits<decltype(size_)>::max() &&
+         "large enough size storage");
+  auto Where = AttributeKeyPool.find(s);
+  if (Where == AttributeKeyPool.end()) {
+    AttributeKey Key(s);
+    AttributeKeyPool.insert({Key.value(), Key});
+    return Key;
+  } else {
+    return Where->second;
+  }
+}

llvm/lib/IR/AutoUpgrade.cpp

   if (!R.match(DL, &Groups))
     return std::string(DL);
 
   return (Groups[1] + AddrSpaces + Groups[3]).str();
 }
 
 void llvm::UpgradeAttributes(AttrBuilder &B) {
   StringRef FramePointer;
-  if (B.contains("no-frame-pointer-elim")) {
+  if (B.contains(NoFramePointerElimAttr)) {
     // The value can be "true" or "false".
     for (const auto &I : B.td_attrs())
-      if (I.first == "no-frame-pointer-elim")
+      if (I.first == NoFramePointerElimAttr)
         FramePointer = I.second == "true" ? "all" : "none";
-    B.removeAttribute("no-frame-pointer-elim");
+    B.removeAttribute(NoFramePointerElimAttr);
   }
-  if (B.contains("no-frame-pointer-elim-non-leaf")) {
+  if (B.contains(NoFramePointerElimNonLeafAttr)) {
     // The value is ignored. "no-frame-pointer-elim"="true" takes priority.
     if (FramePointer != "all")
       FramePointer = "non-leaf";
-    B.removeAttribute("no-frame-pointer-elim-non-leaf");
+    B.removeAttribute(NoFramePointerElimNonLeafAttr);
   }
   if (!FramePointer.empty())
-    B.addAttribute("frame-pointer", FramePointer);
+    B.addAttribute(FramePointerAttr, FramePointer);
 
-  if (B.contains("null-pointer-is-valid")) {
+  if (B.contains(NullPointerIsValidAttr)) {
     // The value can be "true" or "false".
     bool NullPointerIsValid = false;
     for (const auto &I : B.td_attrs())
-      if (I.first == "null-pointer-is-valid")
+      if (I.first == NullPointerIsValidAttr)
         NullPointerIsValid = I.second == "true";
-    B.removeAttribute("null-pointer-is-valid");
+    B.removeAttribute(NullPointerIsValidAttr);
     if (NullPointerIsValid)
       B.addAttribute(Attribute::NullPointerIsValid);
   }
 }

llvm/lib/IR/Core.cpp

 LLVMTypeRef LLVMGetTypeAttributeValue(LLVMAttributeRef A) {
   auto Attr = unwrap(A);
   return wrap(Attr.getValueAsType());
 }
 
 LLVMAttributeRef LLVMCreateStringAttribute(LLVMContextRef C,
                                            const char *K, unsigned KLength,
                                            const char *V, unsigned VLength) {
-  return wrap(Attribute::get(*unwrap(C), StringRef(K, KLength),
+  return wrap(Attribute::get(*unwrap(C),
+                             AttributeKey::Create(StringRef(K, KLength)),
                              StringRef(V, VLength)));
 }
 
 const char *LLVMGetStringAttributeKind(LLVMAttributeRef A,
                                        unsigned *Length) {
-  auto S = unwrap(A).getKindAsString();
+  auto S = unwrap(A).getKindAsKey().value();
   *Length = S.size();
   return S.data();
 }
 
 const char *LLVMGetStringAttributeValue(LLVMAttributeRef A,
                                         unsigned *Length) {
   auto S = unwrap(A).getValueAsString();
   *Length = S.size();

                                              unsigned KindID) {
   return wrap(unwrap<Function>(F)->getAttributeAtIndex(
       Idx, (Attribute::AttrKind)KindID));
 }
 
 LLVMAttributeRef LLVMGetStringAttributeAtIndex(LLVMValueRef F,
                                                LLVMAttributeIndex Idx,
                                                const char *K, unsigned KLen) {
-  return wrap(
-      unwrap<Function>(F)->getAttributeAtIndex(Idx, StringRef(K, KLen)));
+  return wrap(unwrap<Function>(F)->getAttributeAtIndex(
+      Idx, AttributeKey::Create(StringRef(K, KLen))));
 }
 
 void LLVMRemoveEnumAttributeAtIndex(LLVMValueRef F, LLVMAttributeIndex Idx,
                                     unsigned KindID) {
   unwrap<Function>(F)->removeAttributeAtIndex(Idx, (Attribute::AttrKind)KindID);
 }
 
 void LLVMRemoveStringAttributeAtIndex(LLVMValueRef F, LLVMAttributeIndex Idx,
                                       const char *K, unsigned KLen) {
-  unwrap<Function>(F)->removeAttributeAtIndex(Idx, StringRef(K, KLen));
+  unwrap<Function>(F)->removeAttributeAtIndex(
+      Idx, AttributeKey::Create(StringRef(K, KLen)));
 }
 
 void LLVMAddTargetDependentFunctionAttr(LLVMValueRef Fn, const char *A,
                                         const char *V) {
   Function *Func = unwrap<Function>(Fn);
-  Attribute Attr = Attribute::get(Func->getContext(), A, V);
+  Attribute Attr =
+      Attribute::get(Func->getContext(), AttributeKey::Create(A), V);
   Func->addFnAttr(Attr);
 }
 
 /*--.. Operations on parameters ............................................--*/
 
 unsigned LLVMCountParams(LLVMValueRef FnRef) {
   // This function is strictly redundant to
   //   LLVMCountParamTypes(LLVMGetElementType(LLVMTypeOf(FnRef)))

                                               unsigned KindID) {
   return wrap(unwrap<CallBase>(C)->getAttributeAtIndex(
       Idx, (Attribute::AttrKind)KindID));
 }
 
 LLVMAttributeRef LLVMGetCallSiteStringAttribute(LLVMValueRef C,
                                                 LLVMAttributeIndex Idx,
                                                 const char *K, unsigned KLen) {
-  return wrap(
-      unwrap<CallBase>(C)->getAttributeAtIndex(Idx, StringRef(K, KLen)));
+  return wrap(unwrap<CallBase>(C)->getAttributeAtIndex(
+      Idx, AttributeKey::Create(StringRef(K, KLen))));
 }
 
 void LLVMRemoveCallSiteEnumAttribute(LLVMValueRef C, LLVMAttributeIndex Idx,
                                      unsigned KindID) {
   unwrap<CallBase>(C)->removeAttributeAtIndex(Idx, (Attribute::AttrKind)KindID);
 }
 
 void LLVMRemoveCallSiteStringAttribute(LLVMValueRef C, LLVMAttributeIndex Idx,
                                        const char *K, unsigned KLen) {
-  unwrap<CallBase>(C)->removeAttributeAtIndex(Idx, StringRef(K, KLen));
+  unwrap<CallBase>(C)->removeAttributeAtIndex(
+      Idx, AttributeKey::Create(StringRef(K, KLen)));
 }
 
 LLVMValueRef LLVMGetCalledValue(LLVMValueRef Instr) {
   return wrap(unwrap<CallBase>(Instr)->getCalledOperand());
 }
 
 LLVMTypeRef LLVMGetCalledFunctionType(LLVMValueRef Instr) {
   return wrap(unwrap<CallBase>(Instr)->getFunctionType());

llvm/lib/IR/DiagnosticInfo.cpp

 void OptimizationRemarkAnalysisAliasing::anchor() {}
 
 void llvm::diagnoseDontCall(const CallInst &CI) {
   auto *F = CI.getCalledFunction();
   if (!F)
     return;
 
   for (int i = 0; i != 2; ++i) {
-    auto AttrName = i == 0 ? "dontcall-error" : "dontcall-warn";
+    AttributeKey AttrName = i == 0 ? DontcallErrorAttr : DontcallWarnAttr;
     auto Sev = i == 0 ? DS_Error : DS_Warning;
 
     if (F->hasFnAttribute(AttrName)) {
       unsigned LocCookie = 0;
       auto A = F->getFnAttribute(AttrName);
       if (MDNode *MD = CI.getMetadata("srcloc"))
         LocCookie =
             mdconst::extract<ConstantInt>(MD->getOperand(0))->getZExtValue();

llvm/lib/IR/Function.cpp

Function *Function::createWithDefaultAttr(FunctionType *Ty,
   AttrBuilder B;
   if (M->getUwtable())
     B.addAttribute(Attribute::UWTable);
   switch (M->getFramePointer()) {
   case FramePointerKind::None:
     // 0 ("none") is the default.
     break;
   case FramePointerKind::NonLeaf:
-    B.addAttribute("frame-pointer", "non-leaf");
+    B.addAttribute(FramePointerAttr, "non-leaf");
     break;
   case FramePointerKind::All:
-    B.addAttribute("frame-pointer", "all");
+    B.addAttribute(FramePointerAttr, "all");
     break;
   }
   F->addFnAttrs(B);
   return F;
 }
 
 void Function::removeFromParent() {
   getParent()->getFunctionList().remove(getIterator());

 void Function::addAttributeAtIndex(unsigned i, Attribute Attr) {
   AttributeSets = AttributeSets.addAttributeAtIndex(getContext(), i, Attr);
 }
 
 void Function::addFnAttr(Attribute::AttrKind Kind) {
   AttributeSets = AttributeSets.addFnAttribute(getContext(), Kind);
 }
 
-void Function::addFnAttr(StringRef Kind, StringRef Val) {
+void Function::addFnAttr(AttributeKey Kind, StringRef Val) {
   AttributeSets = AttributeSets.addFnAttribute(getContext(), Kind, Val);
 }
 
 void Function::addFnAttr(Attribute Attr) {
   AttributeSets = AttributeSets.addFnAttribute(getContext(), Attr);
 }
 
 void Function::addFnAttrs(const AttrBuilder &Attrs) {

 void Function::addParamAttrs(unsigned ArgNo, const AttrBuilder &Attrs) {
   AttributeSets = AttributeSets.addParamAttributes(getContext(), ArgNo, Attrs);
 }
 
 void Function::removeAttributeAtIndex(unsigned i, Attribute::AttrKind Kind) {
   AttributeSets = AttributeSets.removeAttributeAtIndex(getContext(), i, Kind);
 }
 
-void Function::removeAttributeAtIndex(unsigned i, StringRef Kind) {
+void Function::removeAttributeAtIndex(unsigned i, AttributeKey Kind) {
   AttributeSets = AttributeSets.removeAttributeAtIndex(getContext(), i, Kind);
 }
 
 void Function::removeFnAttr(Attribute::AttrKind Kind) {
   AttributeSets = AttributeSets.removeFnAttribute(getContext(), Kind);
 }
 
-void Function::removeFnAttr(StringRef Kind) {
+void Function::removeFnAttr(AttributeKey Kind) {
   AttributeSets = AttributeSets.removeFnAttribute(getContext(), Kind);
 }
 
 void Function::removeFnAttrs(const AttrBuilder &Attrs) {
   AttributeSets = AttributeSets.removeFnAttributes(getContext(), Attrs);
 }
 
 void Function::removeRetAttr(Attribute::AttrKind Kind) {
   AttributeSets = AttributeSets.removeRetAttribute(getContext(), Kind);
 }
 
-void Function::removeRetAttr(StringRef Kind) {
+void Function::removeRetAttr(AttributeKey Kind) {
   AttributeSets = AttributeSets.removeRetAttribute(getContext(), Kind);
 }
 
 void Function::removeRetAttrs(const AttrBuilder &Attrs) {
   AttributeSets = AttributeSets.removeRetAttributes(getContext(), Attrs);
 }
 
 void Function::removeParamAttr(unsigned ArgNo, Attribute::AttrKind Kind) {
   AttributeSets = AttributeSets.removeParamAttribute(getContext(), ArgNo, Kind);
 }
 
-void Function::removeParamAttr(unsigned ArgNo, StringRef Kind) {
+void Function::removeParamAttr(unsigned ArgNo, AttributeKey Kind) {
   AttributeSets = AttributeSets.removeParamAttribute(getContext(), ArgNo, Kind);
 }
 
 void Function::removeParamAttrs(unsigned ArgNo, const AttrBuilder &Attrs) {
   AttributeSets =
       AttributeSets.removeParamAttributes(getContext(), ArgNo, Attrs);
 }
 
 void Function::addDereferenceableParamAttr(unsigned ArgNo, uint64_t Bytes) {
   AttributeSets =
       AttributeSets.addDereferenceableParamAttr(getContext(), ArgNo, Bytes);
 }
 
 bool Function::hasFnAttribute(Attribute::AttrKind Kind) const {
   return AttributeSets.hasFnAttr(Kind);
 }
 
-bool Function::hasFnAttribute(StringRef Kind) const {
+bool Function::hasFnAttribute(AttributeKey Kind) const {
   return AttributeSets.hasFnAttr(Kind);
 }
 
 bool Function::hasRetAttribute(Attribute::AttrKind Kind) const {
   return AttributeSets.hasRetAttr(Kind);
 }
 
 bool Function::hasParamAttribute(unsigned ArgNo,
                                  Attribute::AttrKind Kind) const {
   return AttributeSets.hasParamAttr(ArgNo, Kind);
 }
 
 Attribute Function::getAttributeAtIndex(unsigned i,
                                         Attribute::AttrKind Kind) const {
   return AttributeSets.getAttributeAtIndex(i, Kind);
 }
 
-Attribute Function::getAttributeAtIndex(unsigned i, StringRef Kind) const {
+Attribute Function::getAttributeAtIndex(unsigned i, AttributeKey Kind) const {
   return AttributeSets.getAttributeAtIndex(i, Kind);
 }
 
 Attribute Function::getFnAttribute(Attribute::AttrKind Kind) const {
   return AttributeSets.getFnAttr(Kind);
 }
 
-Attribute Function::getFnAttribute(StringRef Kind) const {
+Attribute Function::getFnAttribute(AttributeKey Kind) const {
   return AttributeSets.getFnAttr(Kind);
 }
 
 /// gets the specified attribute from the list of attributes.
 Attribute Function::getParamAttribute(unsigned ArgNo,
                                       Attribute::AttrKind Kind) const {
   return AttributeSets.getParamAttr(ArgNo, Kind);
 }
 
 void Function::addDereferenceableOrNullParamAttr(unsigned ArgNo,
                                                  uint64_t Bytes) {
   AttributeSets = AttributeSets.addDereferenceableOrNullParamAttr(getContext(),
                                                                   ArgNo, Bytes);
 }
 
 DenormalMode Function::getDenormalMode(const fltSemantics &FPType) const {
   if (&FPType == &APFloat::IEEEsingle()) {
-    Attribute Attr = getFnAttribute("denormal-fp-math-f32");
+    Attribute Attr = getFnAttribute(DenormalFPMathf32Attr);
     StringRef Val = Attr.getValueAsString();
     if (!Val.empty())
       return parseDenormalFPAttribute(Val);
 
     // If the f32 variant of the attribute isn't specified, try to use the
     // generic one.
   }
 
-  Attribute Attr = getFnAttribute("denormal-fp-math");
+  Attribute Attr = getFnAttribute(DenormalFPMathAttr);
   return parseDenormalFPAttribute(Attr.getValueAsString());
 }
 
 const std::string &Function::getGC() const {
   assert(hasGC() && "Function has no collector");
   return getContext().getGC(*this);
 }
 

llvm/lib/IR/Instructions.cpp

if (CE->getOpcode() == BitCast)
       V = CE->getOperand(0);
 
   if (auto *F = dyn_cast<Function>(V))
     return F->getAttributes().hasFnAttr(Kind);
 
   return false;
 }
 
-bool CallBase::hasFnAttrOnCalledFunction(StringRef Kind) const {
+bool CallBase::hasFnAttrOnCalledFunction(AttributeKey Kind) const {
   Value *V = getCalledOperand();
   if (auto *CE = dyn_cast<ConstantExpr>(V))
     if (CE->getOpcode() == BitCast)
       V = CE->getOperand(0);
 
   if (auto *F = dyn_cast<Function>(V))
     return F->getAttributes().hasFnAttr(Kind);
 

llvm/lib/IR/Statepoint.cpp

 
 #include "llvm/IR/Statepoint.h"
 
 #include "llvm/IR/Function.h"
 
 using namespace llvm;
 
 bool llvm::isStatepointDirectiveAttr(Attribute Attr) {
-  return Attr.hasAttribute("statepoint-id") ||
-         Attr.hasAttribute("statepoint-num-patch-bytes");
+  return Attr.hasAttribute(StatepointIdAttr) ||
+         Attr.hasAttribute(StatepointNumPatchBytesAttr);
 }
 
 StatepointDirectives
 llvm::parseStatepointDirectivesFromAttrs(AttributeList AS) {
   StatepointDirectives Result;
 
-  Attribute AttrID = AS.getFnAttr("statepoint-id");
+  Attribute AttrID = AS.getFnAttr(StatepointIdAttr);
   uint64_t StatepointID;
   if (AttrID.isStringAttribute())
     if (!AttrID.getValueAsString().getAsInteger(10, StatepointID))
       Result.StatepointID = StatepointID;
 
   uint32_t NumPatchBytes;
-  Attribute AttrNumPatchBytes = AS.getFnAttr("statepoint-num-patch-bytes");
+  Attribute AttrNumPatchBytes = AS.getFnAttr(StatepointNumPatchBytesAttr);
   if (AttrNumPatchBytes.isStringAttribute())
     if (!AttrNumPatchBytes.getValueAsString().getAsInteger(10, NumPatchBytes))
       Result.NumPatchBytes = NumPatchBytes;
 
   return Result;
 }

llvm/lib/IR/Verifier.cpp

#include "llvm/IR/Metadata.def"
 
   void verifySwiftErrorCall(CallBase &Call, const Value *SwiftErrorVal);
   void verifySwiftErrorValue(const Value *SwiftErrorVal);
   void verifyTailCCMustTailAttrs(AttrBuilder Attrs, StringRef Context);
   void verifyMustTailCall(CallInst &CI);
   bool verifyAttributeCount(AttributeList Attrs, unsigned Params);
   void verifyAttributeTypes(AttributeSet Attrs, const Value *V);
   void verifyParameterAttrs(AttributeSet Attrs, Type *Ty, const Value *V);
-  void checkUnsignedBaseTenFuncAttr(AttributeList Attrs, StringRef Attr,
+  void checkUnsignedBaseTenFuncAttr(AttributeList Attrs, AttributeKey Attr,
                                     const Value *V);
   void verifyFunctionAttrs(FunctionType *FT, AttributeList Attrs,
                            const Value *V, bool IsIntrinsic);
   void verifyFunctionMetadata(ArrayRef<std::pair<unsigned, MDNode *>> MDs);
   template <typename T>
   void verifyODRTypeAsScopeOperand(const MDNode &MD, T * = nullptr);
 
   void visitConstantExprsRecursively(const Constant *EntryC);

 
 void Verifier::verifyAttributeTypes(AttributeSet Attrs, const Value *V) {
   for (Attribute A : Attrs) {
 
     if (A.isStringAttribute()) {
 #define GET_ATTR_NAMES
 #define ATTRIBUTE_ENUM(ENUM_NAME, DISPLAY_NAME)
 #define ATTRIBUTE_STRBOOL(ENUM_NAME, DISPLAY_NAME)                             \
-  if (A.getKindAsString() == #DISPLAY_NAME) {                                  \
+  if (A.getKindAsKey().value() == #DISPLAY_NAME) {                             \
     auto V = A.getValueAsString();                                             \
     if (!(V.empty() || V == "true" || V == "false"))                           \
       CheckFailed("invalid value for '" #DISPLAY_NAME "' attribute: " + V +    \
                   "");                                                         \
   }
 
 #include "llvm/IR/Attributes.inc"
       continue;
if (!PTy->isOpaque()) {
       if (Attrs.hasAttribute(Attribute::ElementType)) {
         Assert(Attrs.getElementType() == PTy->getElementType(),
                "Attribute 'elementtype' type does not match parameter!", V);
       }
     }
   }
 }
 
-void Verifier::checkUnsignedBaseTenFuncAttr(AttributeList Attrs, StringRef Attr,
-                                            const Value *V) {
+void Verifier::checkUnsignedBaseTenFuncAttr(AttributeList Attrs,
+                                            AttributeKey Attr, const Value *V) {
   if (Attrs.hasFnAttr(Attr)) {
     StringRef S = Attrs.getFnAttr(Attr).getValueAsString();
     unsigned N;
     if (S.getAsInteger(10, N))
-      CheckFailed("\"" + Attr + "\" takes an unsigned integer: " + S, V);
+      CheckFailed("\"" + Attr.value() + "\" takes an unsigned integer: " + S,
+                  V);
   }
 }
 
 // Check parameter attributes against a function type.
 // The value V is printed in error messages.
 void Verifier::verifyFunctionAttrs(FunctionType *FT, AttributeList Attrs,
                                    const Value *V, bool IsIntrinsic) {
   if (Attrs.isEmpty())
void Verifier::verifyFunctionAttrs(FunctionType *FT, AttributeList Attrs,
   if (Attrs.hasFnAttr(Attribute::VScaleRange)) {
     std::pair<unsigned, unsigned> Args =
         Attrs.getFnAttrs().getVScaleRangeArgs();
 
     if (Args.first > Args.second && Args.second != 0)
       CheckFailed("'vscale_range' minimum cannot be greater than maximum", V);
   }
 
-  if (Attrs.hasFnAttr("frame-pointer")) {
-    StringRef FP = Attrs.getFnAttr("frame-pointer").getValueAsString();
+  if (Attrs.hasFnAttr(FramePointerAttr)) {
+    StringRef FP = Attrs.getFnAttr(FramePointerAttr).getValueAsString();
     if (FP != "all" && FP != "non-leaf" && FP != "none")
       CheckFailed("invalid value for 'frame-pointer' attribute: " + FP, V);
   }
 
-  checkUnsignedBaseTenFuncAttr(Attrs, "patchable-function-prefix", V);
-  checkUnsignedBaseTenFuncAttr(Attrs, "patchable-function-entry", V);
-  checkUnsignedBaseTenFuncAttr(Attrs, "warn-stack-size", V);
+  checkUnsignedBaseTenFuncAttr(Attrs, PatchableFunctionPrefixAttr, V);
+  checkUnsignedBaseTenFuncAttr(Attrs, PatchableFunctionEntryAttr, V);
+  checkUnsignedBaseTenFuncAttr(Attrs, WarnStackSizeAttr, V);
 }
 
 void Verifier::verifyFunctionMetadata(
     ArrayRef<std::pair<unsigned, MDNode *>> MDs) {
   for (const auto &Pair : MDs) {
     if (Pair.first == LLVMContext::MD_prof) {
       MDNode *MD = Pair.second;
       Assert(MD->getNumOperands() >= 2,

llvm/lib/Target/TargetLoweringObjectFile.cpp

 MCSection *TargetLoweringObjectFile::SectionForGlobal(
     const GlobalObject *GO, SectionKind Kind, const TargetMachine &TM) const {
   // Select section name.
   if (GO->hasSection())
     return getExplicitSectionGlobal(GO, Kind, TM);
 
   if (auto *GVar = dyn_cast<GlobalVariable>(GO)) {
     auto Attrs = GVar->getAttributes();
-    if ((Attrs.hasAttribute("bss-section") && Kind.isBSS()) ||
-        (Attrs.hasAttribute("data-section") && Kind.isData()) ||
-        (Attrs.hasAttribute("relro-section") && Kind.isReadOnlyWithRel()) ||
-        (Attrs.hasAttribute("rodata-section") && Kind.isReadOnly()))  {
-       return getExplicitSectionGlobal(GO, Kind, TM);
+    if ((Attrs.hasAttribute(BssSectionAttr) && Kind.isBSS()) ||
+        (Attrs.hasAttribute(DataSectionAttr) && Kind.isData()) ||
+        (Attrs.hasAttribute(RelroSectionAttr) && Kind.isReadOnlyWithRel()) ||
+        (Attrs.hasAttribute(RodataSectionAttr) && Kind.isReadOnly())) {
     }
   }
 
   if (auto *F = dyn_cast<Function>(GO)) {
-    if (F->hasFnAttribute("implicit-section-name"))
+    if (F->hasFnAttribute(ImplicitSectionNameAttr))
       return getExplicitSectionGlobal(GO, Kind, TM);
   }
 
   // Use default section depending on the 'type' of global
   return SelectSectionForGlobal(GO, Kind, TM);
 }
 
 /// This method computes the appropriate section to emit the specified global

llvm/lib/Target/TargetMachine.cpp

 // b) these target options should be passed only on the function
 //    and not on the TargetMachine (via TargetOptions) at all.
 void TargetMachine::resetTargetOptions(const Function &F) const {
 #define RESET_OPTION(X, Y)                                              \
   do {                                                                  \
     Options.X = F.getFnAttribute(Y).getValueAsBool();     \
   } while (0)
 
-  RESET_OPTION(UnsafeFPMath, "unsafe-fp-math");
-  RESET_OPTION(NoInfsFPMath, "no-infs-fp-math");
-  RESET_OPTION(NoNaNsFPMath, "no-nans-fp-math");
-  RESET_OPTION(NoSignedZerosFPMath, "no-signed-zeros-fp-math");
+  RESET_OPTION(UnsafeFPMath, UnsafeFPMathAttr);
+  RESET_OPTION(NoInfsFPMath, NoInfsFPMathAttr);
+  RESET_OPTION(NoNaNsFPMath, NoNansFPMathAttr);
+  RESET_OPTION(NoSignedZerosFPMath, NoSignedZerosFPMathAttr);
 }
 
 /// Returns the code generation relocation model. The choices are static, PIC,
 /// and dynamic-no-pic.
 Reloc::Model TargetMachine::getRelocationModel() const { return RM; }
 
 /// Returns the code model. The choices are small, kernel, medium, large, and
 /// target default.

llvm/lib/Target/X86/X86DynAllocaExpander.cpp

bool X86DynAllocaExpander::runOnMachineFunction(MachineFunction &MF) {
   MRI = &MF.getRegInfo();
   STI = &MF.getSubtarget<X86Subtarget>();
   TII = STI->getInstrInfo();
   TRI = STI->getRegisterInfo();
   StackPtr = TRI->getStackRegister();
   SlotSize = TRI->getSlotSize();
 
   StackProbeSize = 4096;
-  if (MF.getFunction().hasFnAttribute("stack-probe-size")) {
+  if (MF.getFunction().hasFnAttribute(StackProbeSizeAttr)) {
     MF.getFunction()
-        .getFnAttribute("stack-probe-size")
+        .getFnAttribute(StackProbeSizeAttr)
         .getValueAsString()
         .getAsInteger(0, StackProbeSize);
   }
-  NoStackArgProbe = MF.getFunction().hasFnAttribute("no-stack-arg-probe");
+  NoStackArgProbe = MF.getFunction().hasFnAttribute(NoStackArgProbeAttr);
   if (NoStackArgProbe)
     StackProbeSize = INT64_MAX;
 
   LoweringMap Lowerings;
   computeLowerings(MF, Lowerings);
   for (auto &P : Lowerings)
     lower(P.first, P.second);
 
   return true;
 }

llvm/lib/Target/X86/X86FastISel.cpp

   bool IsWin64        = Subtarget->isCallingConvWin64(CC);
 
   // Call / invoke instructions with NoCfCheck attribute require special
   // handling.
   if (CB && CB->doesNoCfCheck())
     return false;
 
   // Functions with no_caller_saved_registers that need special handling.
-  if ((CB && isa<CallInst>(CB) && CB->hasFnAttr("no_caller_saved_registers")))
+  if ((CB && isa<CallInst>(CB) && CB->hasFnAttr(NoCallerSavedRegistersAttr)))
     return false;
 
   // Functions with no_callee_saved_registers that need special handling.
-  if ((CB && CB->hasFnAttr("no_callee_saved_registers")))
+  if ((CB && CB->hasFnAttr(NoCalleeSavedRegistersAttr)))
     return false;
 
   // Functions using thunks for indirect calls need to use SDISel.
   if (Subtarget->useIndirectThunkCalls())
     return false;
 
   // Handle only C, fastcc, and webkit_js calling conventions for now.
   switch (CC) {

llvm/lib/Target/X86/X86FrameLowering.cpp

 // If we're forcing a stack realignment we can't rely on just the frame
 // info, we need to know the ABI stack alignment as well in case we
 // have a call out.  Otherwise just make sure we have some alignment - we'll
 // go with the minimum SlotSize.
 uint64_t X86FrameLowering::calculateMaxStackAlign(const MachineFunction &MF) const {
   const MachineFrameInfo &MFI = MF.getFrameInfo();
   Align MaxAlign = MFI.getMaxAlign(); // Desired stack alignment.
   Align StackAlign = getStackAlign();
-  if (MF.getFunction().hasFnAttribute("stackrealign")) {
+  if (MF.getFunction().hasFnAttribute(StackrealignAttr)) {
     if (MFI.hasCalls())
       MaxAlign = (StackAlign > MaxAlign) ? StackAlign : MaxAlign;
     else if (MaxAlign < SlotSize)
       MaxAlign = Align(SlotSize);
   }
   return MaxAlign.value();
 }
 

llvm/lib/Target/X86/X86ISelDAGToDAG.cpp

public:
 
     StringRef getPassName() const override {
       return "X86 DAG->DAG Instruction Selection";
     }
 
     bool runOnMachineFunction(MachineFunction &MF) override {
       // Reset the subtarget each time through.
       Subtarget = &MF.getSubtarget<X86Subtarget>();
-      IndirectTlsSegRefs = MF.getFunction().hasFnAttribute(
-                             "indirect-tls-seg-refs");
+      IndirectTlsSegRefs =
+          MF.getFunction().hasFnAttribute(IndirectTLSSegRefsAttr);
 
       // OptFor[Min]Size are used in pattern predicates that isel is matching.
       OptForMinSize = MF.getFunction().hasMinSize();
       assert((!OptForMinSize || MF.getFunction().hasOptSize()) &&
              "OptForMinSize implies OptForSize");
 
       SelectionDAGISel::runOnMachineFunction(MF);
       return true;

llvm/lib/Target/X86/X86ISelLowering.cpp

                                const SDLoc &dl, SelectionDAG &DAG) const {
   MachineFunction &MF = DAG.getMachineFunction();
   X86MachineFunctionInfo *FuncInfo = MF.getInfo<X86MachineFunctionInfo>();
 
   // In some cases we need to disable registers from the default CSR list.
   // For example, when they are used for argument passing.
   bool ShouldDisableCalleeSavedRegister =
       CallConv == CallingConv::X86_RegCall ||
-      MF.getFunction().hasFnAttribute("no_caller_saved_registers");
+      MF.getFunction().hasFnAttribute(NoCallerSavedRegistersAttr);
 
   if (CallConv == CallingConv::X86_INTR && !Outs.empty())
     report_fatal_error("X86 interrupts may not return any value");
 
   SmallVector<CCValAssign, 16> RVLocs;
   CCState CCInfo(CallConv, isVarArg, MF, RVLocs, *DAG.getContext());
   CCInfo.AnalyzeReturn(Outs, RetCC_X86);
 

       // how far this slot is from the bottom (since they allocate just enough
       // space to accommodate holding this slot at the correct offset).
       int PSPSymFI = MFI.CreateStackObject(8, Align(8), /*isSpillSlot=*/false);
       EHInfo->PSPSymFrameIdx = PSPSymFI;
     }
   }
 
   if (CallConv == CallingConv::X86_RegCall ||
-      F.hasFnAttribute("no_caller_saved_registers")) {
+      F.hasFnAttribute(NoCallerSavedRegistersAttr)) {
     MachineRegisterInfo &MRI = MF.getRegInfo();
     for (std::pair<Register, Register> Pair : MRI.liveins())
       MRI.disableCalleeSavedRegister(Pair.first);
   }
 
   return Chain;
 }
 

   MachineFunction &MF = DAG.getMachineFunction();
   bool Is64Bit        = Subtarget.is64Bit();
   bool IsWin64        = Subtarget.isCallingConvWin64(CallConv);
   bool IsSibcall      = false;
   bool IsGuaranteeTCO = MF.getTarget().Options.GuaranteedTailCallOpt ||
       CallConv == CallingConv::Tail || CallConv == CallingConv::SwiftTail;
   bool IsCalleePopSRet = !IsGuaranteeTCO && hasCalleePopSRet(Outs, Subtarget);
   X86MachineFunctionInfo *X86Info = MF.getInfo<X86MachineFunctionInfo>();
-  bool HasNCSR = (CB && isa<CallInst>(CB) &&
-                  CB->hasFnAttr("no_caller_saved_registers"));
+  bool HasNCSR =
+      (CB && isa<CallInst>(CB) && CB->hasFnAttr(NoCallerSavedRegistersAttr));
   bool HasNoCfCheck = (CB && CB->doesNoCfCheck());
   bool IsIndirectCall = (CB && isa<CallInst>(CB) && CB->isIndirectCall());
   const Module *M = MF.getMMI().getModule();
   Metadata *IsCFProtectionSupported = M->getModuleFlag("cf-protection-branch");
 
   MachineFunction::CallSiteInfo CSInfo;
   if (CallConv == CallingConv::X86_INTR)
     report_fatal_error("X86 interrupts may not be called directly");

     auto AdaptedCC = CallConv;
     // If HasNCSR is asserted (attribute NoCallerSavedRegisters exists),
     // use X86_INTR calling convention because it has the same CSR mask
     // (same preserved registers).
     if (HasNCSR)
       AdaptedCC = (CallingConv::ID)CallingConv::X86_INTR;
     // If NoCalleeSavedRegisters is requested, than use GHC since it happens
     // to use the CSR_NoRegs_RegMask.
-    if (CB && CB->hasFnAttr("no_callee_saved_registers"))
+    if (CB && CB->hasFnAttr(NoCalleeSavedRegistersAttr))
       AdaptedCC = (CallingConv::ID)CallingConv::GHC;
     return RegInfo->getCallPreservedMask(MF, AdaptedCC);
   }();
   assert(Mask && "Missing call preserved mask for calling convention");
 
   // If this is an invoke in a 32-bit function using a funclet-based
   // personality, assume the function clobbers all registers. If an exception
   // is thrown, the runtime will not restore CSRs.

   return !getStackProbeSymbolName(MF).empty();
 }
 
 /// Returns true if stack probing through inline assembly is requested.
 bool X86TargetLowering::hasInlineStackProbe(MachineFunction &MF) const {
 
   // No inline stack probe for Windows, they have their own mechanism.
   if (Subtarget.isOSWindows() ||
-      MF.getFunction().hasFnAttribute("no-stack-arg-probe"))
+      MF.getFunction().hasFnAttribute(NoStackArgProbeAttr))
     return false;
 
   // If the function specifically requests inline stack probes, emit them.
-  if (MF.getFunction().hasFnAttribute("probe-stack"))
-    return MF.getFunction().getFnAttribute("probe-stack").getValueAsString() ==
+  if (MF.getFunction().hasFnAttribute(ProbeStackAttr))
+    return MF.getFunction().getFnAttribute(ProbeStackAttr).getValueAsString() ==
            "inline-asm";
 
   return false;
 }
 
 /// Returns the name of the symbol used to emit stack probes or the empty
 /// string if not applicable.
 StringRef
 X86TargetLowering::getStackProbeSymbolName(MachineFunction &MF) const {
   // Inline Stack probes disable stack probe call
   if (hasInlineStackProbe(MF))
     return "";
 
   // If the function specifically requests stack probes, emit them.
-  if (MF.getFunction().hasFnAttribute("probe-stack"))
-    return MF.getFunction().getFnAttribute("probe-stack").getValueAsString();
+  if (MF.getFunction().hasFnAttribute(ProbeStackAttr))
+    return MF.getFunction().getFnAttribute(ProbeStackAttr).getValueAsString();
 
   // Generally, if we aren't on Windows, the platform ABI does not include
   // support for stack probes, so don't emit them.
   if (!Subtarget.isOSWindows() || Subtarget.isTargetMachO() ||
-      MF.getFunction().hasFnAttribute("no-stack-arg-probe"))
+      MF.getFunction().hasFnAttribute(NoStackArgProbeAttr))
     return "";
 
   // We need a stack probe to conform to the Windows ABI. Choose the right
   // symbol.
   if (Subtarget.is64Bit())
     return Subtarget.isTargetCygMing() ? "___chkstk_ms" : "__chkstk";
   return Subtarget.isTargetCygMing() ? "_alloca" : "_chkstk";
 }
 
 unsigned
 X86TargetLowering::getStackProbeSize(MachineFunction &MF) const {
   // The default stack probe size is 4096 if the function has no stackprobesize
   // attribute.
   unsigned StackProbeSize = 4096;
   const Function &Fn = MF.getFunction();
-  if (Fn.hasFnAttribute("stack-probe-size"))
-    Fn.getFnAttribute("stack-probe-size")
+  if (Fn.hasFnAttribute(StackProbeSizeAttr))
+    Fn.getFnAttribute(StackProbeSizeAttr)
         .getValueAsString()
         .getAsInteger(0, StackProbeSize);
   return StackProbeSize;
 }
 
 Align X86TargetLowering::getPrefLoopAlignment(MachineLoop *ML) const {
   if (ML->isInnermost() &&
       ExperimentalPrefInnermostLoopAlignment.getNumOccurrences())
     return Align(1ULL << ExperimentalPrefInnermostLoopAlignment);
   return TargetLowering::getPrefLoopAlignment();
 }

llvm/lib/Target/X86/X86MCInstLower.cpp

 }
 
 void X86AsmPrinter::LowerPATCHABLE_FUNCTION_ENTER(const MachineInstr &MI,
                                                   X86MCInstLower &MCIL) {
 
   NoAutoPaddingScope NoPadScope(*OutStreamer);
 
   const Function &F = MF->getFunction();
-  if (F.hasFnAttribute("patchable-function-entry")) {
+  if (F.hasFnAttribute(PatchableFunctionEntryAttr)) {
     unsigned Num;
-    if (F.getFnAttribute("patchable-function-entry")
+    if (F.getFnAttribute(PatchableFunctionEntryAttr)
             .getValueAsString()
             .getAsInteger(10, Num))
       return;
     emitX86Nops(*OutStreamer, Num, Subtarget);
     return;
   }
   // We want to emit the following pattern:
   //

llvm/lib/Target/X86/X86RegisterInfo.cpp

X86RegisterInfo::getCalleeSavedRegs(const MachineFunction *MF) const {
   bool HasAVX = Subtarget.hasAVX();
   bool HasAVX512 = Subtarget.hasAVX512();
   bool CallsEHReturn = MF->callsEHReturn();
 
   CallingConv::ID CC = F.getCallingConv();
 
   // If attribute NoCallerSavedRegisters exists then we set X86_INTR calling
   // convention because it has the CSR list.
-  if (MF->getFunction().hasFnAttribute("no_caller_saved_registers"))
+  if (MF->getFunction().hasFnAttribute(NoCallerSavedRegistersAttr))
     CC = CallingConv::X86_INTR;
 
   // If atribute specified, override the CSRs normally specified by the
   // calling convention and use the empty set instead.
-  if (MF->getFunction().hasFnAttribute("no_callee_saved_registers"))
+  if (MF->getFunction().hasFnAttribute(NoCalleeSavedRegistersAttr))
     return CSR_NoRegs_SaveList;
 
   switch (CC) {
   case CallingConv::GHC:
   case CallingConv::HiPE:
     return CSR_NoRegs_SaveList;
   case CallingConv::AnyReg:
     if (HasAVX)

llvm/lib/Target/X86/X86TargetMachine.cpp

X86TargetMachine::X86TargetMachine(const Target &T, const Triple &TT,
 
   initAsmInfo();
 }
 
 X86TargetMachine::~X86TargetMachine() = default;
 
 const X86Subtarget *
 X86TargetMachine::getSubtargetImpl(const Function &F) const {
-  Attribute CPUAttr = F.getFnAttribute("target-cpu");
-  Attribute TuneAttr = F.getFnAttribute("tune-cpu");
-  Attribute FSAttr = F.getFnAttribute("target-features");
+  Attribute CPUAttr = F.getFnAttribute(TargetCPUAttr);
+  Attribute TuneAttr = F.getFnAttribute(TuneCPUAttr);
+  Attribute FSAttr = F.getFnAttribute(TargetFeaturesAttr);
 
   StringRef CPU =
       CPUAttr.isValid() ? CPUAttr.getValueAsString() : (StringRef)TargetCPU;
   StringRef TuneCPU =
       TuneAttr.isValid() ? TuneAttr.getValueAsString() : (StringRef)CPU;
   StringRef FS =
       FSAttr.isValid() ? FSAttr.getValueAsString() : (StringRef)TargetFS;
 
   SmallString<512> Key;
   // The additions here are ordered so that the definitely short strings are
   // added first so we won't exceed the small size. We append the
   // much longer FS string at the end so that we only heap allocate at most
   // one time.
 
   // Extract prefer-vector-width attribute.
   unsigned PreferVectorWidthOverride = 0;
-  Attribute PreferVecWidthAttr = F.getFnAttribute("prefer-vector-width");
+  Attribute PreferVecWidthAttr = F.getFnAttribute(PreferVectorWidthAttr);
   if (PreferVecWidthAttr.isValid()) {
     StringRef Val = PreferVecWidthAttr.getValueAsString();
     unsigned Width;
     if (!Val.getAsInteger(0, Width)) {
       Key += 'p';
       Key += Val;
       PreferVectorWidthOverride = Width;
     }
   }
 
   // Extract min-legal-vector-width attribute.
   unsigned RequiredVectorWidth = UINT32_MAX;
-  Attribute MinLegalVecWidthAttr = F.getFnAttribute("min-legal-vector-width");
+  Attribute MinLegalVecWidthAttr = F.getFnAttribute(MinLegalVectorWidthAttr);
   if (MinLegalVecWidthAttr.isValid()) {
     StringRef Val = MinLegalVecWidthAttr.getValueAsString();
     unsigned Width;
     if (!Val.getAsInteger(0, Width)) {
       Key += 'm';
       Key += Val;
       RequiredVectorWidth = Width;
     }
   }
 
   // Add CPU to the Key.
   Key += CPU;
 
   // Add tune CPU to the Key.
   Key += TuneCPU;
 
   // Keep track of the start of the feature portion of the string.
   unsigned FSStart = Key.size();
 
   // FIXME: This is related to the code below to reset the target options,
   // we need to know whether or not the soft float flag is set on the
   // function before we can generate a subtarget. We also need to use
   // it as a key for the subtarget since that can be the only difference
   // between two functions.
-  bool SoftFloat = F.getFnAttribute("use-soft-float").getValueAsBool();
+  bool SoftFloat = F.getFnAttribute(UseSoftFloatAttr).getValueAsBool();
   // If the soft float attribute is set on the function turn on the soft float
   // subtarget feature.
   if (SoftFloat)
     Key += FS.empty() ? "+soft-float" : "+soft-float,";
 
   Key += FS;
 
   // We may have added +soft-float to the features so move the StringRef to

llvm/lib/Target/X86/X86WinEHState.cpp

Function *WinEHStatePass::generateLSDAInEAXThunk(Function *ParentFunc) {
   Call->addParamAttr(0, Attribute::InReg);
   Builder.CreateRet(Call);
   return Trampoline;
 }
 
 void WinEHStatePass::linkExceptionRegistration(IRBuilder<> &Builder,
                                                Function *Handler) {
   // Emit the .safeseh directive for this function.
-  Handler->addFnAttr("safeseh");
+  Handler->addFnAttr(SafesehAttr);
 
   Type *LinkTy = getEHLinkRegistrationType();
   // Handler = Handler
   Value *HandlerI8 = Builder.CreateBitCast(Handler, Builder.getInt8PtrTy());
   Builder.CreateStore(HandlerI8, Builder.CreateStructGEP(LinkTy, Link, 1));
   // Next = [fs:00]
   Constant *FSZero =
       Constant::getNullValue(LinkTy->getPointerTo()->getPointerTo(257));

llvm/lib/Transforms/CFGuard/CFGuard.cpp

bool CFGuard::runOnFunction(Function &F) {
 
   // Iterate over the instructions to find all indirect call/invoke/callbr
   // instructions. Make a separate list of pointers to indirect
   // call/invoke/callbr instructions because the original instructions will be
   // deleted as the checks are added.
   for (BasicBlock &BB : F.getBasicBlockList()) {
     for (Instruction &I : BB.getInstList()) {
       auto *CB = dyn_cast<CallBase>(&I);
-      if (CB && CB->isIndirectCall() && !CB->hasFnAttr("guard_nocf")) {
+      if (CB && CB->isIndirectCall() && !CB->hasFnAttr(GuardNocfAttr)) {
         IndirectCalls.push_back(CB);
         CFGuardCounter++;
       }
     }
   }
 
   // If no checks are needed, return early.
   if (IndirectCalls.empty()) {

llvm/lib/Transforms/Coroutines/CoroEarly.cpp

if (auto *CB = dyn_cast<CallBase>(&I)) {
       case Intrinsic::coro_noop:
         lowerCoroNoop(cast<IntrinsicInst>(&I));
         break;
       case Intrinsic::coro_id:
         // Mark a function that comes out of the frontend that has a coro.id
         // with a coroutine attribute.
         if (auto *CII = cast<CoroIdInst>(&I)) {
           if (CII->getInfo().isPreSplit()) {
-            F.addFnAttr(CORO_PRESPLIT_ATTR, UNPREPARED_FOR_SPLIT);
+            F.addFnAttr(CoroutinePresplitAttr, UNPREPARED_FOR_SPLIT);
             setCannotDuplicate(CII);
             CII->setCoroutineSelf();
             CoroId = cast<CoroIdInst>(&I);
           }
         }
         break;
       case Intrinsic::coro_id_retcon:
       case Intrinsic::coro_id_retcon_once:
       case Intrinsic::coro_id_async:
-        F.addFnAttr(CORO_PRESPLIT_ATTR, PREPARED_FOR_SPLIT);
+        F.addFnAttr(CoroutinePresplitAttr, PREPARED_FOR_SPLIT);
         break;
       case Intrinsic::coro_resume:
         lowerResumeOrDestroy(*CB, CoroSubFnInst::ResumeIndex);
         break;
       case Intrinsic::coro_destroy:
         lowerResumeOrDestroy(*CB, CoroSubFnInst::DestroyIndex);
         break;
       case Intrinsic::coro_promise:

llvm/lib/Transforms/Coroutines/CoroElide.cpp

struct CoroElideLegacy : FunctionPass {
   }
 
   bool runOnFunction(Function &F) override {
     if (!L)
       return false;
 
     bool Changed = false;
 
-    if (F.hasFnAttribute(CORO_PRESPLIT_ATTR))
+    if (F.hasFnAttribute(CoroutinePresplitAttr))
       Changed = replaceDevirtTrigger(F);
 
     L->CoroIds.clear();
     L->collectPostSplitCoroIds(&F);
     // If we did not find any coro.id, there is nothing to do.
     if (L->CoroIds.empty())
       return Changed;
 

llvm/lib/Transforms/Coroutines/CoroInternal.h

 // twice. First, it lets it go through complete IPO optimization pipeline as a
 // single function. It forces restart of the pipeline by inserting an indirect
 // call to an empty function "coro.devirt.trigger" which is devirtualized by
 // CoroElide pass that triggers a restart of the pipeline by CGPassManager.
 // When CoroSplit pass sees the same coroutine the second time, it splits it up,
 // adds coroutine subfunctions to the SCC to be processed by IPO pipeline.
 // Async lowering similarily triggers a restart of the pipeline after it has
 // split the coroutine.
-#define CORO_PRESPLIT_ATTR "coroutine.presplit"
 #define UNPREPARED_FOR_SPLIT "0"
 #define PREPARED_FOR_SPLIT "1"
 #define ASYNC_RESTART_AFTER_SPLIT "2"
 
 #define CORO_DEVIRT_TRIGGER_FN "coro.devirt.trigger"
 
 namespace coro {
 

llvm/lib/Transforms/Coroutines/CoroSplit.cpp

                             bool MarkForAsyncRestart = false) {
   Module &M = *F.getParent();
   LLVMContext &Context = F.getContext();
 #ifndef NDEBUG
   Function *DevirtFn = M.getFunction(CORO_DEVIRT_TRIGGER_FN);
   assert(DevirtFn && "coro.devirt.trigger function not found");
 #endif
 
-  F.addFnAttr(CORO_PRESPLIT_ATTR, MarkForAsyncRestart
+  F.addFnAttr(CoroutinePresplitAttr, MarkForAsyncRestart
                                          ? ASYNC_RESTART_AFTER_SPLIT
                                          : PREPARED_FOR_SPLIT);
 
   // Insert an indirect call sequence that will be devirtualized by CoroElide
   // pass:
   //    %0 = call i8* @llvm.coro.subfn.addr(i8* null, i8 -1)
   //    %1 = bitcast i8* %0 to void(i8*)*
   //    call void %1(i8* null)
   coro::LowererBase Lowerer(M);
   Instruction *InsertPt =

   // Check for uses of llvm.coro.prepare.retcon/async.
   SmallVector<Function *, 2> PrepareFns;
   addPrepareFunction(M, PrepareFns, "llvm.coro.prepare.retcon");
   addPrepareFunction(M, PrepareFns, "llvm.coro.prepare.async");
 
   // Find coroutines for processing.
   SmallVector<LazyCallGraph::Node *, 4> Coroutines;
   for (LazyCallGraph::Node &N : C)
-    if (N.getFunction().hasFnAttribute(CORO_PRESPLIT_ATTR))
+    if (N.getFunction().hasFnAttribute(CoroutinePresplitAttr))
       Coroutines.push_back(&N);
 
   if (Coroutines.empty() && PrepareFns.empty())
     return PreservedAnalyses::all();
 
   if (Coroutines.empty()) {
     for (auto *PrepareFn : PrepareFns) {
       replaceAllPrepares(PrepareFn, CG, C);
     }
   }
 
   // Split all the coroutines.
   for (LazyCallGraph::Node *N : Coroutines) {
     Function &F = N->getFunction();
-    LLVM_DEBUG(dbgs() << "CoroSplit: Processing coroutine '" << F.getName()
-                      << "' state: "
-                      << F.getFnAttribute(CORO_PRESPLIT_ATTR).getValueAsString()
-                      << "\n");
-    F.removeFnAttr(CORO_PRESPLIT_ATTR);
+    LLVM_DEBUG(
+        dbgs() << "CoroSplit: Processing coroutine '" << F.getName()
+               << "' state: "
+               << F.getFnAttribute(CoroutinePresplitAttr).getValueAsString()
+               << "\n");
+    F.removeFnAttr(CoroutinePresplitAttr);
 
     SmallVector<Function *, 4> Clones;
     const coro::Shape Shape = splitCoroutine(F, Clones, ReuseFrameSlot);
     updateCallGraphAfterCoroutineSplit(*N, Shape, Clones, C, CG, AM, UR, FAM);
 
     if (!Shape.CoroSuspends.empty()) {
       // Run the CGSCC pipeline on the original and newly split functions.
       UR.CWorklist.insert(&C);

     auto &M = SCC.getCallGraph().getModule();
     addPrepareFunction(M, PrepareFns, "llvm.coro.prepare.retcon");
     addPrepareFunction(M, PrepareFns, "llvm.coro.prepare.async");
 
     // Find coroutines for processing.
     SmallVector<Function *, 4> Coroutines;
     for (CallGraphNode *CGN : SCC)
       if (auto *F = CGN->getFunction())
-        if (F->hasFnAttribute(CORO_PRESPLIT_ATTR))
+        if (F->hasFnAttribute(CoroutinePresplitAttr))
           Coroutines.push_back(F);
 
     if (Coroutines.empty() && PrepareFns.empty())
       return false;
 
     CallGraph &CG = getAnalysis<CallGraphWrapperPass>().getCallGraph();
 
     if (Coroutines.empty()) {
       bool Changed = false;
       for (auto *PrepareFn : PrepareFns)
         Changed |= replaceAllPrepares(PrepareFn, CG);
       return Changed;
     }
 
     createDevirtTriggerFunc(CG, SCC);
 
     // Split all the coroutines.
     for (Function *F : Coroutines) {
-      Attribute Attr = F->getFnAttribute(CORO_PRESPLIT_ATTR);
+      Attribute Attr = F->getFnAttribute(CoroutinePresplitAttr);
       StringRef Value = Attr.getValueAsString();
       LLVM_DEBUG(dbgs() << "CoroSplit: Processing coroutine '" << F->getName()
                         << "' state: " << Value << "\n");
       // Async lowering marks coroutines to trigger a restart of the pipeline
       // after it has split them.
       if (Value == ASYNC_RESTART_AFTER_SPLIT) {
-        F->removeFnAttr(CORO_PRESPLIT_ATTR);
+        F->removeFnAttr(CoroutinePresplitAttr);
         continue;
       }
       if (Value == UNPREPARED_FOR_SPLIT) {
         prepareForSplit(*F, CG);
         continue;
       }
-      F->removeFnAttr(CORO_PRESPLIT_ATTR);
+      F->removeFnAttr(CoroutinePresplitAttr);
 
       SmallVector<Function *, 4> Clones;
       const coro::Shape Shape = splitCoroutine(*F, Clones, ReuseFrameSlot);
       updateCallGraphAfterCoroutineSplit(*F, Shape, Clones, CG, SCC);
       if (Shape.ABI == coro::ABI::Async) {
         // Restart SCC passes.
         // Mark function for CoroElide pass. It will devirtualize causing a
         // restart of the SCC pipeline.

llvm/lib/Transforms/IPO/Attributor.cpp

if (Attr.isEnumAttribute()) {
     if (Attrs.hasAttributeAtIndex(AttrIdx, Kind))
       if (!ForceReplace &&
           isEqualOrWorse(Attr, Attrs.getAttributeAtIndex(AttrIdx, Kind)))
         return false;
     Attrs = Attrs.addAttributeAtIndex(Ctx, AttrIdx, Attr);
     return true;
   }
   if (Attr.isStringAttribute()) {
-    StringRef Kind = Attr.getKindAsString();
+    AttributeKey Kind = Attr.getKindAsKey();
     if (Attrs.hasAttributeAtIndex(AttrIdx, Kind))
       if (!ForceReplace &&
           isEqualOrWorse(Attr, Attrs.getAttributeAtIndex(AttrIdx, Kind)))
         return false;
     Attrs = Attrs.addAttributeAtIndex(Ctx, AttrIdx, Attr);
     return true;
   }
   if (Attr.isIntAttribute()) {

llvm/lib/Transforms/IPO/AttributorAttributes.cpp

                        SmallVectorImpl<Attribute> &Attrs) const override {
     if (!isAssumedNoCaptureMaybeReturned())
       return;
 
     if (isArgumentPosition()) {
       if (isAssumedNoCapture())
         Attrs.emplace_back(Attribute::get(Ctx, Attribute::NoCapture));
       else if (ManifestInternal)
-        Attrs.emplace_back(Attribute::get(Ctx, "no-capture-maybe-returned"));
+        Attrs.emplace_back(Attribute::get(Ctx, NoCaptureMaybeReturned));
     }
   }
 
   /// Set the NOT_CAPTURED_IN_MEM and NOT_CAPTURED_IN_RET bits in \p Known
   /// depending on the ability of the function associated with \p IRP to capture
   /// state in memory and through "returning/throwing", respectively.
   static void determineFunctionCaptureCapabilities(const IRPosition &IRP,
                                                    const Function &F,

llvm/lib/Transforms/IPO/CrossDSOCFI.cpp

void CrossDSOCFI::buildCFICheck(Module &M) {
   Function *F = cast<Function>(C.getCallee());
   // Take over the existing function. The frontend emits a weak stub so that the
   // linker knows about the symbol; this pass replaces the function body.
   F->deleteBody();
   F->setAlignment(Align(4096));
 
   Triple T(M.getTargetTriple());
   if (T.isARM() || T.isThumb())
-    F->addFnAttr("target-features", "+thumb-mode");
+    F->addFnAttr(TargetFeaturesAttr, "+thumb-mode");
 
   auto args = F->arg_begin();
   Value &CallSiteTypeId = *(args++);
   CallSiteTypeId.setName("CallSiteTypeId");
   Value &Addr = *(args++);
   Addr.setName("Addr");
   Value &CFICheckFailData = *(args++);
   CFICheckFailData.setName("CFICheckFailData");

llvm/lib/Transforms/IPO/FunctionImport.cpp

 }
 
 // Internalize values that we marked with specific attribute
 // in processGlobalForThinLTO.
 static void internalizeGVsAfterImport(Module &M) {
   for (auto &GV : M.globals())
     // Skip GVs which have been converted to declarations
     // by dropDeadSymbols.
-    if (!GV.isDeclaration() && GV.hasAttribute("thinlto-internalize")) {
+    if (!GV.isDeclaration() && GV.hasAttribute(ThinltoInternalizeAttr)) {
       GV.setLinkage(GlobalValue::InternalLinkage);
       GV.setVisibility(GlobalValue::DefaultVisibility);
     }
 }
 
 // Automatically import functions in Module \p DestModule based on the summaries
 // index.
 Expected<bool> FunctionImporter::importFunctions(

llvm/lib/Transforms/IPO/GlobalOpt.cpp

 #include "llvm/Support/raw_ostream.h"
 #include "llvm/Transforms/IPO.h"
 #include "llvm/Transforms/Utils/CtorUtils.h"
 #include "llvm/Transforms/Utils/Evaluator.h"
 #include "llvm/Transforms/Utils/GlobalStatus.h"
 #include "llvm/Transforms/Utils/Local.h"
 #include <cassert>
 #include <cstdint>
+#include <map>
 #include <utility>
 #include <vector>
 
 using namespace llvm;
 
 #define DEBUG_TYPE "globalopt"
 
 STATISTIC(NumMarked    , "Number of globals marked constant");

llvm/lib/Transforms/IPO/LowerTypeTests.cpp

 
 bool lowertypetests::isJumpTableCanonical(Function *F) {
   if (F->isDeclarationForLinker())
     return false;
   auto *CI = mdconst::extract_or_null<ConstantInt>(
       F->getParent()->getModuleFlag("CFI Canonical Jump Tables"));
   if (!CI || CI->getZExtValue() != 0)
     return true;
-  return F->hasFnAttribute("cfi-canonical-jump-table");
+  return F->hasFnAttribute(CfiCanonicalJumpTablesAttr);
 }
 
 namespace {
 
 struct ByteArrayInfo {
   std::set<uint64_t> Bits;
   uint64_t BitSize;
   GlobalVariable *ByteArray;
Constant *Target = ConstantExpr::getSelect(
       ConstantExpr::getICmp(CmpInst::ICMP_NE, F,
                             Constant::getNullValue(F->getType())),
       JT, Constant::getNullValue(F->getType()));
   PlaceholderFn->replaceAllUsesWith(Target);
   PlaceholderFn->eraseFromParent();
 }
 
 static bool isThumbFunction(Function *F, Triple::ArchType ModuleArch) {
-  Attribute TFAttr = F->getFnAttribute("target-features");
+  Attribute TFAttr = F->getFnAttribute(TargetFeaturesAttr);
   if (TFAttr.isValid()) {
     SmallVector<StringRef, 6> Features;
     TFAttr.getValueAsString().split(Features, ',');
     for (StringRef Feature : Features) {
       if (Feature == "-thumb-mode")
         return false;
       else if (Feature == "+thumb-mode")
         return true;
void LowerTypeTestsModule::createJumpTable(
   F->setAlignment(Align(getJumpTableEntrySize()));
   // Skip prologue.
   // Disabled on win32 due to https://llvm.org/bugs/show_bug.cgi?id=28641#c3.
   // Luckily, this function does not get any prologue even without the
   // attribute.
   if (OS != Triple::Win32)
     F->addFnAttr(Attribute::Naked);
   if (JumpTableArch == Triple::arm)
-    F->addFnAttr("target-features", "-thumb-mode");
+    F->addFnAttr(TargetFeaturesAttr, "-thumb-mode");
   if (JumpTableArch == Triple::thumb) {
-    F->addFnAttr("target-features", "+thumb-mode");
+    F->addFnAttr(TargetFeaturesAttr, "+thumb-mode");
     // Thumb jump table assembly needs Thumb2. The following attribute is added
     // by Clang for -march=armv7.
-    F->addFnAttr("target-cpu", "cortex-a8");
+    F->addFnAttr(TargetCPUAttr, "cortex-a8");
   }
   if (JumpTableArch == Triple::aarch64) {
-    F->addFnAttr("branch-target-enforcement", "false");
-    F->addFnAttr("sign-return-address", "none");
+    F->addFnAttr(BranchTargetEnforcementAttr, "false");
+    F->addFnAttr(SignReturnAddressAttr, "none");
   }
   // Make sure we don't emit .eh_frame for this function.
   F->addFnAttr(Attribute::NoUnwind);
 
   BasicBlock *BB = BasicBlock::Create(M.getContext(), "entry", F);
   IRBuilder<> IRB(BB);
 
   SmallVector<Type *, 16> ArgTypes;

llvm/lib/Transforms/IPO/OpenMPOpt.cpp

   }
 
   /// Hepler to check if \p I is a call and get the value for it if it is
   /// unique.
   Optional<Value *> getValueForCall(Attributor &A, const Instruction *I,
                                     InternalControlVar &ICV) const {
 
     const auto *CB = dyn_cast<CallBase>(I);
-    if (!CB || CB->hasFnAttr("no_openmp") ||
-        CB->hasFnAttr("no_openmp_routines"))
+    if (!CB || CB->hasFnAttr(NoOpenmp) || CB->hasFnAttr(NoOpenmpRoutines))
       return None;
 
     auto &OMPInfoCache = static_cast<OMPInformationCache &>(A.getInfoCache());
     auto &GetterRFI = OMPInfoCache.RFIs[OMPInfoCache.ICVs[ICV].Getter];
     auto &SetterRFI = OMPInfoCache.RFIs[OMPInfoCache.ICVs[ICV].Setter];
     Function *CalledFunction = CB->getCalledFunction();
 
     // Indirect call, assume ICV changes.

       break;
     case OMPRTL___kmpc_is_generic_main_thread_id:
       Changed |= foldIsGenericMainThread(A);
       break;
     case OMPRTL___kmpc_parallel_level:
       Changed |= foldParallelLevel(A);
       break;
     case OMPRTL___kmpc_get_hardware_num_threads_in_block:
-      Changed = Changed | foldKernelFnAttribute(A, "omp_target_thread_limit");
+      Changed = Changed | foldKernelFnAttribute(A, OmpTargetThreadLimitAttr);
       break;
     case OMPRTL___kmpc_get_hardware_num_blocks:
-      Changed = Changed | foldKernelFnAttribute(A, "omp_target_num_teams");
+      Changed = Changed | foldKernelFnAttribute(A, OmpTargetNumTeamsAttr);
       break;
     default:
       llvm_unreachable("Unhandled OpenMP runtime function!");
     }
 
     return Changed;
   }
 

       assert(KnownSPMDCount == 0 && AssumedSPMDCount == 0 &&
              "Expected only non-SPMD kernels!");
       SimplifiedValue = ConstantInt::get(Type::getInt8Ty(Ctx), 0);
     }
     return SimplifiedValue == SimplifiedValueBefore ? ChangeStatus::UNCHANGED
                                                     : ChangeStatus::CHANGED;
   }
 
-  ChangeStatus foldKernelFnAttribute(Attributor &A, llvm::StringRef Attr) {
+  ChangeStatus foldKernelFnAttribute(Attributor &A, AttributeKey Attr) {
     // Specialize only if all the calls agree with the attribute constant value
     int32_t CurrentAttrValue = -1;
     Optional<Value *> SimplifiedValueBefore = SimplifiedValue;
 
     auto &CallerKernelInfoAA = A.getAAFor<AAKernelInfo>(
         *this, IRPosition::function(*getAnchorScope()), DepClassTy::REQUIRED);
 
     if (!CallerKernelInfoAA.ReachingKernelEntries.isValidState())

llvm/lib/Transforms/IPO/SampleProfile.cpp

bool SampleProfileLoader::tryPromoteAndInlineCandidate(
   const char *Reason = "Callee function not available";
   // R->getValue() != &F is to prevent promoting a recursive call.
   // If it is a recursive call, we do not inline it as it could bloat
   // the code exponentially. There is way to better handle this, e.g.
   // clone the caller first, and inline the cloned caller if it is
   // recursive. As llvm does not inline recursive calls, we will
   // simply ignore it instead of handling it explicitly.
   if (!R->getValue()->isDeclaration() && R->getValue()->getSubprogram() &&
-      R->getValue()->hasFnAttribute("use-sample-profile") &&
+      R->getValue()->hasFnAttribute(UseSampleProfileAttr) &&
       R->getValue() != &F && isLegalToPromote(CI, R->getValue(), &Reason)) {
     // For promoted target, set its value with NOMORE_ICP_MAGICNUM count
     // in the value profile metadata so the target won't be promoted again.
     SmallVector<InstrProfValueData, 1> SortedCallTargets = {InstrProfValueData{
         Function::getGUID(R->getValue()->getName()), NOMORE_ICP_MAGICNUM}};
     updateIDTMetaData(CI, SortedCallTargets, 0);
 
     auto *DI = &pgo::promoteIndirectCall(

 ///
 /// \returns True if there is any inline happened.
 bool SampleProfileLoader::inlineHotFunctions(
     Function &F, DenseSet<GlobalValue::GUID> &InlinedGUIDs) {
   // ProfAccForSymsInList is used in callsiteIsHot. The assertion makes sure
   // Profile symbol list is ignored when profile-sample-accurate is on.
   assert((!ProfAccForSymsInList ||
           (!ProfileSampleAccurate &&
-           !F.hasFnAttribute("profile-sample-accurate"))) &&
+           !F.hasFnAttribute(ProfileSampleAccurateAttr))) &&
          "ProfAccForSymsInList should be false when profile-sample-accurate "
          "is enabled");
 
   DenseMap<CallBase *, const FunctionSamples *> LocalNotInlinedCallSites;
   bool Changed = false;
   bool LocalChanged = true;
   while (LocalChanged) {
     LocalChanged = false;

 bool SampleProfileLoader::inlineHotFunctionsWithPriority(
     Function &F, DenseSet<GlobalValue::GUID> &InlinedGUIDs) {
   assert(ProfileIsCS && "Prioritiy based inliner only works with CSSPGO now");
 
   // ProfAccForSymsInList is used in callsiteIsHot. The assertion makes sure
   // Profile symbol list is ignored when profile-sample-accurate is on.
   assert((!ProfAccForSymsInList ||
           (!ProfileSampleAccurate &&
-           !F.hasFnAttribute("profile-sample-accurate"))) &&
+           !F.hasFnAttribute(ProfileSampleAccurateAttr))) &&
          "ProfAccForSymsInList should be false when profile-sample-accurate "
          "is enabled");
 
   // Populating worklist with initial call sites from root inliner, along
   // with call site weights.
   CandidateQueue CQueue;
   InlineCandidate NewCandidate;
   for (auto &BB : F) {
SampleProfileLoader::buildProfiledCallGraph(CallGraph &CG) {
   else
     ProfiledCG = std::make_unique<ProfiledCallGraph>(Reader->getProfiles());
 
   // Add all functions into the profiled call graph even if they are not in
   // the profile. This makes sure functions missing from the profile still
   // gets a chance to be processed.
   for (auto &Node : CG) {
     const auto *F = Node.first;
-    if (!F || F->isDeclaration() || !F->hasFnAttribute("use-sample-profile"))
+    if (!F || F->isDeclaration() || !F->hasFnAttribute(UseSampleProfileAttr))
       continue;
     ProfiledCG->addProfiledFunction(FunctionSamples::getCanonicalFnName(*F));
   }
 
   return ProfiledCG;
 }
 
 std::vector<Function *>
if (ProfileMergeInlinee) {
       // because the profile for a function may be used for the profile
       // annotation of its outline copy before the profile merging of its
       // non-inlined inline instances, and that is not the way how
       // ProfileMergeInlinee is supposed to work.
       ProfileMergeInlinee = false;
     }
 
     for (Function &F : M)
-      if (!F.isDeclaration() && F.hasFnAttribute("use-sample-profile"))
+      if (!F.isDeclaration() && F.hasFnAttribute(UseSampleProfileAttr))
         FunctionOrderList.push_back(&F);
     return FunctionOrderList;
   }
 
   assert(&CG->getModule() == &M);
 
   if (UseProfiledCallGraph ||
       (ProfileIsCS && !UseProfiledCallGraph.getNumOccurrences())) {
if (UseProfiledCallGraph ||
     // static call edges are not so important when they don't correspond to a
     // context in the profile.
 
     std::unique_ptr<ProfiledCallGraph> ProfiledCG = buildProfiledCallGraph(*CG);
     scc_iterator<ProfiledCallGraph *> CGI = scc_begin(ProfiledCG.get());
     while (!CGI.isAtEnd()) {
       for (ProfiledCallGraphNode *Node : *CGI) {
         Function *F = SymbolMap.lookup(Node->Name);
-        if (F && !F->isDeclaration() && F->hasFnAttribute("use-sample-profile"))
+        if (F && !F->isDeclaration() && F->hasFnAttribute(UseSampleProfileAttr))
           FunctionOrderList.push_back(F);
       }
       ++CGI;
     }
   } else {
     scc_iterator<CallGraph *> CGI = scc_begin(CG);
     while (!CGI.isAtEnd()) {
       for (CallGraphNode *Node : *CGI) {
         auto *F = Node->getFunction();
-        if (F && !F->isDeclaration() && F->hasFnAttribute("use-sample-profile"))
+        if (F && !F->isDeclaration() && F->hasFnAttribute(UseSampleProfileAttr))
           FunctionOrderList.push_back(F);
       }
       ++CGI;
     }
   }
 
   LLVM_DEBUG({
     dbgs() << "Function processing order:\n";
bool SampleProfileLoader::runOnFunction(Function &F, ModuleAnalysisManager *AM) {
   DILocation2SampleMap.clear();
   // By default the entry count is initialized to -1, which will be treated
   // conservatively by getEntryCount as the same as unknown (None). This is
   // to avoid newly added code to be treated as cold. If we have samples
   // this will be overwritten in emitAnnotations.
   uint64_t initialEntryCount = -1;
 
   ProfAccForSymsInList = ProfileAccurateForSymsInList && PSL;
-  if (ProfileSampleAccurate || F.hasFnAttribute("profile-sample-accurate")) {
+  if (ProfileSampleAccurate || F.hasFnAttribute(ProfileSampleAccurateAttr)) {
     // initialize all the function entry counts to 0. It means all the
     // functions without profile will be regarded as cold.
     initialEntryCount = 0;
     // profile-sample-accurate is a user assertion which has a higher precedence
     // than symbol list. When profile-sample-accurate is on, ignore symbol list.
     ProfAccForSymsInList = false;
   }
   CoverageTracker.setProfAccForSymsInList(ProfAccForSymsInList);

llvm/lib/Transforms/IPO/WholeProgramDevirt.cpp

     if (CSInfo.isExported())
       IsExported = true;
     if (CSInfo.AllCallSitesDevirted)
       return;
     for (auto &&VCallSite : CSInfo.CallSites) {
       CallBase &CB = VCallSite.CB;
 
       // Jump tables are only profitable if the retpoline mitigation is enabled.
-      Attribute FSAttr = CB.getCaller()->getFnAttribute("target-features");
+      Attribute FSAttr = CB.getCaller()->getFnAttribute(TargetFeaturesAttr);
       if (!FSAttr.isValid() ||
           !FSAttr.getValueAsString().contains("+retpoline"))
         continue;
 
       if (RemarksEnabled)
         VCallSite.emitRemark("branch-funnel",
                              JT->stripPointerCasts()->getName(), OREGetter);
 

llvm/lib/Transforms/InstCombine/InstCombineCalls.cpp

   auto *Callee =
       dyn_cast<Function>(Call.getCalledOperand()->stripPointerCasts());
   if (!Callee)
     return false;
 
   // If this is a call to a thunk function, don't remove the cast. Thunks are
   // used to transparently forward all incoming parameters and outgoing return
   // values, so it's important to leave the cast in place.
-  if (Callee->hasFnAttribute("thunk"))
+  if (Callee->hasFnAttribute(ThunkAttr))
     return false;
 
   // If this is a musttail call, the callee's prototype must match the caller's
   // prototype with the exception of pointee types. The code below doesn't
   // implement that, so we can't do this transform.
   // TODO: Do the transform if it only requires adding pointer casts.
   if (Call.isMustTailCall())
     return false;

llvm/lib/Transforms/Instrumentation/DataFlowSanitizer.cpp

   Function *NewF = Function::Create(NewFT, NewFLink, F->getAddressSpace(),
                                     NewFName, F->getParent());
   NewF->copyAttributesFrom(F);
   NewF->removeRetAttrs(
       AttributeFuncs::typeIncompatible(NewFT->getReturnType()));
 
   BasicBlock *BB = BasicBlock::Create(*Ctx, "entry", NewF);
   if (F->isVarArg()) {
-    NewF->removeFnAttrs(AttrBuilder().addAttribute("split-stack"));
+    NewF->removeFnAttrs(AttrBuilder().addAttribute(SplitStackAttr));
     CallInst::Create(DFSanVarargWrapperFn,
                      IRBuilder<>(BB).CreateGlobalStringPtr(F->getName()), "",
                      BB);
     new UnreachableInst(*Ctx, BB);
   } else {
     auto ArgIt = pointer_iterator<Argument *>(NewF->arg_begin());
     std::vector<Value *> Args(ArgIt, ArgIt + FT->getNumParams());
 

llvm/lib/Transforms/Instrumentation/MemorySanitizer.cpp

   enum ArgKind { AK_GeneralPurpose, AK_FloatingPoint, AK_Memory };
 
   VarArgAMD64Helper(Function &F, MemorySanitizer &MS,
                     MemorySanitizerVisitor &MSV)
       : F(F), MS(MS), MSV(MSV) {
     AMD64FpEndOffset = AMD64FpEndOffsetSSE;
     for (const auto &Attr : F.getAttributes().getFnAttrs()) {
       if (Attr.isStringAttribute() &&
-          (Attr.getKindAsString() == "target-features")) {
+          (Attr.getKindAsKey().value() == "target-features")) {
         if (Attr.getValueAsString().contains("-sse"))
           AMD64FpEndOffset = AMD64FpEndOffsetNoSSE;
         break;
       }
     }
   }
 
   ArgKind classifyArgument(Value* arg) {

       assert(!ZExt);
       return ShadowExtension::Sign;
     }
     return ShadowExtension::None;
   }
 
   void visitCallBase(CallBase &CB, IRBuilder<> &IRB) override {
     bool IsSoftFloatABI = CB.getCalledFunction()
-                              ->getFnAttribute("use-soft-float")
+                              ->getFnAttribute(UseSoftFloatAttr)
                               .getValueAsBool();
     unsigned GpOffset = SystemZGpOffset;
     unsigned FpOffset = SystemZFpOffset;
     unsigned VrIndex = 0;
     unsigned OverflowOffset = SystemZOverflowOffset;
     const DataLayout &DL = F.getParent()->getDataLayout();
     for (auto ArgIt = CB.arg_begin(), End = CB.arg_end(); ArgIt != End;
          ++ArgIt) {

llvm/lib/Transforms/Instrumentation/ThreadSanitizer.cpp

for (auto Inst : AtomicAccesses) {
       Res |= instrumentAtomic(Inst, DL);
     }
 
   if (ClInstrumentMemIntrinsics && SanitizeFunction)
     for (auto Inst : MemIntrinCalls) {
       Res |= instrumentMemIntrinsic(Inst);
     }
 
-  if (F.hasFnAttribute("sanitize_thread_no_checking_at_run_time")) {
+  if (F.hasFnAttribute(SanitizeThreadNoCheckingAtRunTimeAttr)) {
     assert(!F.hasFnAttribute(Attribute::SanitizeThread));
     if (HasCalls)
       InsertRuntimeIgnores(F);
   }
 
   // Instrument function entry/exit points if there were instrumented accesses.
   if ((Res || HasCalls) && ClInstrumentFuncEntryExit) {
     IRBuilder<> IRB(F.getEntryBlock().getFirstNonPHI());

llvm/lib/Transforms/ObjCARC/ObjCARCOpts.cpp

 static bool isInertARCValue(Value *V, SmallPtrSet<Value *, 1> &VisitedPhis) {
   V = V->stripPointerCasts();
 
   if (IsNullOrUndef(V))
     return true;
 
   // See if this is a global attribute annotated with an 'objc_arc_inert'.
   if (auto *GV = dyn_cast<GlobalVariable>(V))
-    if (GV->hasAttribute("objc_arc_inert"))
+    if (GV->hasAttribute(ObjcArcInertAttr))
       return true;
 
   if (auto PN = dyn_cast<PHINode>(V)) {
     // Ignore this phi if it has already been discovered.
     if (!VisitedPhis.insert(PN).second)
       return true;
     // Look through phis's operands.
     for (Value *Opnd : PN->incoming_values())

llvm/lib/Transforms/Scalar/RewriteStatepointsForGC.cpp

 
     OldI->eraseFromParent();
   }
 };
 
 } // end anonymous namespace
 
 static StringRef getDeoptLowering(CallBase *Call) {
-  const char *DeoptLowering = "deopt-lowering";
-  if (Call->hasFnAttr(DeoptLowering)) {
+  if (Call->hasFnAttr(DeoptLoweringAttr)) {
     // FIXME: Calls have a *really* confusing interface around attributes
     // with values.
     const AttributeList &CSAS = Call->getAttributes();
-    if (CSAS.hasFnAttr(DeoptLowering))
-      return CSAS.getFnAttr(DeoptLowering).getValueAsString();
+    if (CSAS.hasFnAttr(DeoptLoweringAttr))
+      return CSAS.getFnAttr(DeoptLoweringAttr).getValueAsString();
     Function *F = Call->getCalledFunction();
-    assert(F && F->hasFnAttribute(DeoptLowering));
-    return F->getFnAttribute(DeoptLowering).getValueAsString();
+    assert(F && F->hasFnAttribute(DeoptLoweringAttr));
+    return F->getFnAttribute(DeoptLoweringAttr).getValueAsString();
   }
   return "live-through";
 }
 
 static void
 makeStatepointExplicitImpl(CallBase *Call, /* to replace */
                            const SmallVectorImpl<Value *> &BasePtrs,
                            const SmallVectorImpl<Value *> &LiveVariables,

llvm/lib/Transforms/Scalar/TailRecursionElimination.cpp

bool TailRecursionEliminator::processBlock(BasicBlock &BB) {
   return false;
 }
 
 bool TailRecursionEliminator::eliminate(Function &F,
                                         const TargetTransformInfo *TTI,
                                         AliasAnalysis *AA,
                                         OptimizationRemarkEmitter *ORE,
                                         DomTreeUpdater &DTU) {
-  if (F.getFnAttribute("disable-tail-calls").getValueAsBool())
+  if (F.getFnAttribute(DisableTailCallsAttr).getValueAsBool())
     return false;
 
   bool MadeChange = false;
   MadeChange |= markTails(F, ORE);
 
   // If this function is a varargs function, we won't be able to PHI the args
   // right, so don't even try to convert it...
   if (F.getFunctionType()->isVarArg())

llvm/lib/Transforms/Utils/CodeExtractor.cpp

Function *CodeExtractor::constructFunction(const ValueSet &inputs,
   // target independent attributes.
   //  (e.g. If the extracted region contains a call to an x86.sse
   //  instruction we need to make sure that the extracted region has the
   //  "target-features" attribute allowing it to be lowered.
   // FIXME: This should be changed to check to see if a specific
   //           attribute can not be inherited.
   for (const auto &Attr : oldFunction->getAttributes().getFnAttrs()) {
     if (Attr.isStringAttribute()) {
-      if (Attr.getKindAsString() == "thunk")
+      if (Attr.getKindAsKey().value() == "thunk")
         continue;
     } else
       switch (Attr.getKindAsEnum()) {
       // Those attributes cannot be propagated safely. Explicitly list them
       // here so we get a warning if new attributes are added. This list also
       // includes non-function attributes.
       case Attribute::Alignment:
       case Attribute::AllocSize:

llvm/lib/Transforms/Utils/EntryExitInstrumenter.cpp

static void insertCall(Function &CurFn, StringRef Func,
   }
 
   // We only know how to call a fixed set of instrumentation functions, because
   // they all expect different arguments, etc.
   report_fatal_error(Twine("Unknown instrumentation function: '") + Func + "'");
 }
 
 static bool runOnFunction(Function &F, bool PostInlining) {
-  StringRef EntryAttr = PostInlining ? "instrument-function-entry-inlined"
-                                     : "instrument-function-entry";
+  AttributeKey EntryAttr = PostInlining ? InstrumentFunctionEntryInlinedAttr
+                                        : InstrumentFunctionEntryAttr;
 
-  StringRef ExitAttr = PostInlining ? "instrument-function-exit-inlined"
-                                    : "instrument-function-exit";
+  AttributeKey ExitAttr = PostInlining ? InstrumentFunctionExitInlinedAttr
+                                       : InstrumentFunctionExitAttr;
 
   StringRef EntryFunc = F.getFnAttribute(EntryAttr).getValueAsString();
   StringRef ExitFunc = F.getFnAttribute(ExitAttr).getValueAsString();
 
   bool Changed = false;
 
   // If the attribute is specified, insert instrumentation and then "consume"
   // the attribute so that it's not inserted again if the pass should happen to

llvm/lib/Transforms/Utils/FunctionImportUtils.cpp

if (GlobalVariable *V = dyn_cast<GlobalVariable>(&GV)) {
       // can happen in the distributed ThinLTO backend, because the index only
       // contains summaries from the source modules if they are being imported.
       // We might have a non-null VI and get here even in that case if the name
       // matches one in this module (e.g. weak or appending linkage).
       auto *GVS = dyn_cast_or_null<GlobalVarSummary>(
           ImportIndex.findSummaryInModule(VI, M.getModuleIdentifier()));
       if (GVS &&
           (ImportIndex.isReadOnly(GVS) || ImportIndex.isWriteOnly(GVS))) {
-        V->addAttribute("thinlto-internalize");
+        V->addAttribute(ThinltoInternalizeAttr);
         // Objects referenced by writeonly GV initializer should not be
         // promoted, because there is no any kind of read access to them
         // on behalf of this writeonly GV. To avoid promotion we convert
         // GV initializer to 'zeroinitializer'. This effectively drops
         // references in IR module (not in combined index), so we can
         // ignore them when computing import. We do not export references
         // of writeonly object. See computeImportForReferencedGlobals
         if (ImportIndex.isWriteOnly(GVS))

llvm/lib/Transforms/Utils/InlineFunction.cpp

 
   // Cache the inlined-at nodes as they're built so they are reused, without
   // this every instruction's inlined-at chain would become distinct from each
   // other.
   DenseMap<const MDNode *, MDNode *> IANodes;
 
   // Check if we are not generating inline line tables and want to use
   // the call site location instead.
-  bool NoInlineLineTables = Fn->hasFnAttribute("no-inline-line-tables");
+  bool NoInlineLineTables = Fn->hasFnAttribute(NoInlineLineTablesAttr);
 
   for (; FI != Fn->end(); ++FI) {
     for (BasicBlock::iterator BI = FI->begin(), BE = FI->end();
          BI != BE; ++BI) {
       // Loop metadata needs to be updated so that the start and end locs
       // reference inlined-at locations.
       auto updateLoopInfoLoc = [&Ctx, &InlinedAtNode,
                                 &IANodes](Metadata *MD) -> Metadata * {

llvm/lib/Transforms/Utils/Local.cpp

     return DT.dominates(BB, U);
   };
   return ::replaceDominatedUsesWith(From, To, BB, Dominates);
 }
 
 bool llvm::callsGCLeafFunction(const CallBase *Call,
                                const TargetLibraryInfo &TLI) {
   // Check if the function is specifically marked as a gc leaf function.
-  if (Call->hasFnAttr("gc-leaf-function"))
+  if (Call->hasFnAttr(GcLeafFunctionAttr))
     return true;
   if (const Function *F = Call->getCalledFunction()) {
-    if (F->hasFnAttribute("gc-leaf-function"))
+    if (F->hasFnAttribute(GcLeafFunctionAttr))
       return true;
 
     if (auto IID = F->getIntrinsicID()) {
       // Most LLVM intrinsics do not take safepoints.
       return IID != Intrinsic::experimental_gc_statepoint &&
              IID != Intrinsic::experimental_deoptimize &&
              IID != Intrinsic::memcpy_element_unordered_atomic &&
              IID != Intrinsic::memmove_element_unordered_atomic;

llvm/lib/Transforms/Utils/ModuleUtils.cpp

for (const std::string &VariantMapping : VariantMappings) {
     LLVM_DEBUG(dbgs() << "VFABI: adding mapping '" << VariantMapping << "'\n");
     Optional<VFInfo> VI = VFABI::tryDemangleForVFABI(VariantMapping, *M);
     assert(VI.hasValue() && "Cannot add an invalid VFABI name.");
     assert(M->getNamedValue(VI.getValue().VectorName) &&
            "Cannot add variant to attribute: "
            "vector function declaration is missing.");
   }
 #endif
-  CI->addFnAttr(
-      Attribute::get(M->getContext(), MappingsAttrName, Buffer.str()));
+  CI->addFnAttr(Attribute::get(M->getContext(), VectorFunctionAbiVariantAttr,
+                               Buffer.str()));
 }

llvm/lib/Transforms/Utils/SimplifyCFG.cpp

                                 const TargetTransformInfo &TTI) {
   assert(SI->getNumCases() > 1 && "Degenerate switch?");
 
   BasicBlock *BB = SI->getParent();
   Function *Fn = BB->getParent();
   // Only build lookup table when we have a target that supports it or the
   // attribute is not set.
   if (!TTI.shouldBuildLookupTables() ||
-      (Fn->getFnAttribute("no-jump-tables").getValueAsBool()))
+      (Fn->getFnAttribute(NoJumpTablesAttr).getValueAsBool()))
     return false;
 
   // FIXME: If the switch is too sparse for a lookup table, perhaps we could
   // split off a dense part and build a lookup table for that.
 
   // FIXME: This creates arrays of GEPs to constant strings, which means each
   // GEP needs a runtime relocation in PIC code. We should just build one big
   // string and lookup indices into that.

llvm/tools/bugpoint-passes/TestPasses.cpp

 
 private:
   void getAnalysisUsage(AnalysisUsage &AU) const override {
     AU.setPreservesAll();
   }
 
   bool runOnFunction(Function &F) override {
     AttributeSet A = F.getAttributes().getFnAttrs();
-    if (A.hasAttribute("bugpoint-crash"))
+    if (A.hasAttribute(AttributeKey::Create("bugpoint-crash")))
       abort();
     return false;
   }
 };
 } // namespace
 
 char CrashOnFunctionAttribute::ID = 0;
 static RegisterPass<CrashOnFunctionAttribute>

llvm/unittests/IR/InstructionsTest.cpp

TEST_F(ModuleWithFunctionTest, CallInst) {
   // Make sure iteration over a call's arguments works as expected.
   unsigned Idx = 0;
   for (Value *Arg : Call->args()) {
     EXPECT_EQ(FArgTypes[Idx], Arg->getType());
     EXPECT_EQ(Call->getArgOperand(Idx)->getType(), Arg->getType());
     Idx++;
   }
 
-  Call->addRetAttr(Attribute::get(Call->getContext(), "test-str-attr"));
-  EXPECT_TRUE(Call->hasRetAttr("test-str-attr"));
-  EXPECT_FALSE(Call->hasRetAttr("not-on-call"));
+  Call->addRetAttr(Attribute::get(Call->getContext(),
+                                  AttributeKey::Create("test-str-attr")));
+  EXPECT_TRUE(Call->hasRetAttr(AttributeKey::Create("test-str-attr")));
+  EXPECT_FALSE(Call->hasRetAttr(AttributeKey::Create("not-on-call")));
 }
 
 TEST_F(ModuleWithFunctionTest, InvokeInst) {
   BasicBlock *BB1 = BasicBlock::Create(Ctx, "", F);
   BasicBlock *BB2 = BasicBlock::Create(Ctx, "", F);
 
   Value *Args[] = {ConstantInt::get(Type::getInt8Ty(Ctx), 20),
                    ConstantInt::get(Type::getInt32Ty(Ctx), 9999),

llvm/unittests/ProfileData/SampleProfTest.cpp

struct SampleProfTest : ::testing::Test {
   void createFunctionWithSampleProfileElisionPolicy(Module *M,
                                                     const char *Fname,
                                                     StringRef Policy) {
     FunctionType *FnType =
         FunctionType::get(Type::getVoidTy(Context), {}, false);
     auto Inserted = M->getOrInsertFunction(Fname, FnType);
     auto Fcn = cast<Function>(Inserted.getCallee());
     if (Policy != "")
-      Fcn->addFnAttr("sample-profile-suffix-elision-policy", Policy);
+      Fcn->addFnAttr(SampleProfileSuffixElisionPolicyAttr, Policy);
   }
 
   void setupModuleForElisionTest(Module *M, StringRef Policy) {
     createFunctionWithSampleProfileElisionPolicy(M, "foo", Policy);
     createFunctionWithSampleProfileElisionPolicy(M, "foo.bar", Policy);
     createFunctionWithSampleProfileElisionPolicy(M, "foo.llvm.2465", Policy);
   }
 

llvm/unittests/Transforms/Utils/VFABIUtils.cpp

protected:
   SmallVector<std::string, 8> Mappings;
 };
 
 TEST_F(VFABIAttrTest, Write) {
   Mappings.push_back("_ZGVnN8v_g");
   Mappings.push_back("_ZGVnN2v_g(custom_vg)");
   VFABI::setVectorVariantNames(CI, Mappings);
   const StringRef S =
-      CI->getFnAttr("vector-function-abi-variant").getValueAsString();
+      CI->getFnAttr(llvm::AttributeKey::Create("vector-function-abi-variant"))
+          .getValueAsString();
   EXPECT_EQ(S, "_ZGVnN8v_g,_ZGVnN2v_g(custom_vg)");
 }