diff --git a/src/reader/spirv/function.cc b/src/reader/spirv/function.cc index 0767da6d15..46fb806f2b 100644 --- a/src/reader/spirv/function.cc +++ b/src/reader/spirv/function.cc @@ -470,19 +470,19 @@ std::string GetGlslStd450FuncName(uint32_t ext_opcode) { } // Returns the WGSL standard library function intrinsic for the -// given instruction, or semantic::Intrinsic::kNone -semantic::Intrinsic GetIntrinsic(SpvOp opcode) { +// given instruction, or semantic::IntrinsicType::kNone +semantic::IntrinsicType GetIntrinsic(SpvOp opcode) { switch (opcode) { case SpvOpBitCount: - return semantic::Intrinsic::kCountOneBits; + return semantic::IntrinsicType::kCountOneBits; case SpvOpBitReverse: - return semantic::Intrinsic::kReverseBits; + return semantic::IntrinsicType::kReverseBits; case SpvOpDot: - return semantic::Intrinsic::kDot; + return semantic::IntrinsicType::kDot; default: break; } - return semantic::Intrinsic::kNone; + return semantic::IntrinsicType::kNone; } // @param opcode a SPIR-V opcode @@ -3212,7 +3212,7 @@ TypedExpression FunctionEmitter::MaybeEmitCombinatorialValue( } const auto intrinsic = GetIntrinsic(opcode); - if (intrinsic != semantic::Intrinsic::kNone) { + if (intrinsic != semantic::IntrinsicType::kNone) { return MakeIntrinsicCall(inst); } diff --git a/src/semantic/call.h b/src/semantic/call.h index c1459d5209..6351421139 100644 --- a/src/semantic/call.h +++ b/src/semantic/call.h @@ -40,16 +40,16 @@ class IntrinsicCall : public Castable { /// Constructor /// @param return_type the return type of the call /// @param intrinsic the call target intrinsic - IntrinsicCall(type::Type* return_type, Intrinsic intrinsic); + IntrinsicCall(type::Type* return_type, IntrinsicType intrinsic); /// Destructor ~IntrinsicCall() override; /// @returns the target intrinsic for the call - Intrinsic intrinsic() const { return intrinsic_; } + IntrinsicType intrinsic() const { return intrinsic_; } private: - Intrinsic const intrinsic_; + IntrinsicType const intrinsic_; }; /// TextureIntrinsicCall holds semantic information for ast::CallExpression @@ -105,7 +105,7 @@ class TextureIntrinsicCall /// @param intrinsic the call target intrinsic /// @param params the overload parameter info TextureIntrinsicCall(type::Type* return_type, - Intrinsic intrinsic, + IntrinsicType intrinsic, const Parameters& params); /// Destructor diff --git a/src/semantic/intrinsic.cc b/src/semantic/intrinsic.cc index 2ecd238a25..b2eea277e7 100644 --- a/src/semantic/intrinsic.cc +++ b/src/semantic/intrinsic.cc @@ -17,218 +17,221 @@ namespace tint { namespace semantic { -std::ostream& operator<<(std::ostream& out, Intrinsic i) { +std::ostream& operator<<(std::ostream& out, IntrinsicType i) { out << intrinsic::str(i); return out; } namespace intrinsic { -const char* str(Intrinsic i) { +const char* str(IntrinsicType i) { /// The emitted name matches the spelling in the WGSL spec. /// including case. switch (i) { - case Intrinsic::kNone: + case IntrinsicType::kNone: return ""; - case Intrinsic::kAbs: + case IntrinsicType::kAbs: return "abs"; - case Intrinsic::kAcos: + case IntrinsicType::kAcos: return "acos"; - case Intrinsic::kAll: + case IntrinsicType::kAll: return "all"; - case Intrinsic::kAny: + case IntrinsicType::kAny: return "any"; - case Intrinsic::kArrayLength: + case IntrinsicType::kArrayLength: return "arrayLength"; - case Intrinsic::kAsin: + case IntrinsicType::kAsin: return "asin"; - case Intrinsic::kAtan: + case IntrinsicType::kAtan: return "atan"; - case Intrinsic::kAtan2: + case IntrinsicType::kAtan2: return "atan2"; - case Intrinsic::kCeil: + case IntrinsicType::kCeil: return "ceil"; - case Intrinsic::kClamp: + case IntrinsicType::kClamp: return "clamp"; - case Intrinsic::kCos: + case IntrinsicType::kCos: return "cos"; - case Intrinsic::kCosh: + case IntrinsicType::kCosh: return "cosh"; - case Intrinsic::kCountOneBits: + case IntrinsicType::kCountOneBits: return "countOneBits"; - case Intrinsic::kCross: + case IntrinsicType::kCross: return "cross"; - case Intrinsic::kDeterminant: + case IntrinsicType::kDeterminant: return "determinant"; - case Intrinsic::kDistance: + case IntrinsicType::kDistance: return "distance"; - case Intrinsic::kDot: + case IntrinsicType::kDot: return "dot"; - case Intrinsic::kDpdx: + case IntrinsicType::kDpdx: return "dpdx"; - case Intrinsic::kDpdxCoarse: + case IntrinsicType::kDpdxCoarse: return "dpdxCoarse"; - case Intrinsic::kDpdxFine: + case IntrinsicType::kDpdxFine: return "dpdxFine"; - case Intrinsic::kDpdy: + case IntrinsicType::kDpdy: return "dpdy"; - case Intrinsic::kDpdyCoarse: + case IntrinsicType::kDpdyCoarse: return "dpdyCoarse"; - case Intrinsic::kDpdyFine: + case IntrinsicType::kDpdyFine: return "dpdyFine"; - case Intrinsic::kExp: + case IntrinsicType::kExp: return "exp"; - case Intrinsic::kExp2: + case IntrinsicType::kExp2: return "exp2"; - case Intrinsic::kFaceForward: + case IntrinsicType::kFaceForward: return "faceForward"; - case Intrinsic::kFloor: + case IntrinsicType::kFloor: return "floor"; - case Intrinsic::kFma: + case IntrinsicType::kFma: return "fma"; - case Intrinsic::kFract: + case IntrinsicType::kFract: return "fract"; - case Intrinsic::kFrexp: + case IntrinsicType::kFrexp: return "frexp"; - case Intrinsic::kFwidth: + case IntrinsicType::kFwidth: return "fwidth"; - case Intrinsic::kFwidthCoarse: + case IntrinsicType::kFwidthCoarse: return "fwidthCoarse"; - case Intrinsic::kFwidthFine: + case IntrinsicType::kFwidthFine: return "fwidthFine"; - case Intrinsic::kInverseSqrt: + case IntrinsicType::kInverseSqrt: return "inverseSqrt"; - case Intrinsic::kIsFinite: + case IntrinsicType::kIsFinite: return "isFinite"; - case Intrinsic::kIsInf: + case IntrinsicType::kIsInf: return "isInf"; - case Intrinsic::kIsNan: + case IntrinsicType::kIsNan: return "isNan"; - case Intrinsic::kIsNormal: + case IntrinsicType::kIsNormal: return "isNormal"; - case Intrinsic::kLdexp: + case IntrinsicType::kLdexp: return "ldexp"; - case Intrinsic::kLength: + case IntrinsicType::kLength: return "length"; - case Intrinsic::kLog: + case IntrinsicType::kLog: return "log"; - case Intrinsic::kLog2: + case IntrinsicType::kLog2: return "log2"; - case Intrinsic::kMax: + case IntrinsicType::kMax: return "max"; - case Intrinsic::kMin: + case IntrinsicType::kMin: return "min"; - case Intrinsic::kMix: + case IntrinsicType::kMix: return "mix"; - case Intrinsic::kModf: + case IntrinsicType::kModf: return "modf"; - case Intrinsic::kNormalize: + case IntrinsicType::kNormalize: return "normalize"; - case Intrinsic::kPack4x8Snorm: + case IntrinsicType::kPack4x8Snorm: return "pack4x8snorm"; - case Intrinsic::kPack4x8Unorm: + case IntrinsicType::kPack4x8Unorm: return "pack4x8unorm"; - case Intrinsic::kPack2x16Snorm: + case IntrinsicType::kPack2x16Snorm: return "pack2x16snorm"; - case Intrinsic::kPack2x16Unorm: + case IntrinsicType::kPack2x16Unorm: return "pack2x16unorm"; - case Intrinsic::kPack2x16Float: + case IntrinsicType::kPack2x16Float: return "pack2x16float"; - case Intrinsic::kPow: + case IntrinsicType::kPow: return "pow"; - case Intrinsic::kReflect: + case IntrinsicType::kReflect: return "reflect"; - case Intrinsic::kReverseBits: + case IntrinsicType::kReverseBits: return "reverseBits"; - case Intrinsic::kRound: + case IntrinsicType::kRound: return "round"; - case Intrinsic::kSelect: + case IntrinsicType::kSelect: return "select"; - case Intrinsic::kSign: + case IntrinsicType::kSign: return "sign"; - case Intrinsic::kSin: + case IntrinsicType::kSin: return "sin"; - case Intrinsic::kSinh: + case IntrinsicType::kSinh: return "sinh"; - case Intrinsic::kSmoothStep: + case IntrinsicType::kSmoothStep: return "smoothStep"; - case Intrinsic::kSqrt: + case IntrinsicType::kSqrt: return "sqrt"; - case Intrinsic::kStep: + case IntrinsicType::kStep: return "step"; - case Intrinsic::kTan: + case IntrinsicType::kTan: return "tan"; - case Intrinsic::kTanh: + case IntrinsicType::kTanh: return "tanh"; - case Intrinsic::kTextureDimensions: + case IntrinsicType::kTextureDimensions: return "textureDimensions"; - case Intrinsic::kTextureLoad: + case IntrinsicType::kTextureLoad: return "textureLoad"; - case Intrinsic::kTextureNumLayers: + case IntrinsicType::kTextureNumLayers: return "textureNumLayers"; - case Intrinsic::kTextureNumLevels: + case IntrinsicType::kTextureNumLevels: return "textureNumLevels"; - case Intrinsic::kTextureNumSamples: + case IntrinsicType::kTextureNumSamples: return "textureNumSamples"; - case Intrinsic::kTextureSample: + case IntrinsicType::kTextureSample: return "textureSample"; - case Intrinsic::kTextureSampleBias: + case IntrinsicType::kTextureSampleBias: return "textureSampleBias"; - case Intrinsic::kTextureSampleCompare: + case IntrinsicType::kTextureSampleCompare: return "textureSampleCompare"; - case Intrinsic::kTextureSampleGrad: + case IntrinsicType::kTextureSampleGrad: return "textureSampleGrad"; - case Intrinsic::kTextureSampleLevel: + case IntrinsicType::kTextureSampleLevel: return "textureSampleLevel"; - case Intrinsic::kTextureStore: + case IntrinsicType::kTextureStore: return "textureStore"; - case Intrinsic::kTrunc: + case IntrinsicType::kTrunc: return "trunc"; } return ""; } -bool IsCoarseDerivative(Intrinsic i) { - return i == Intrinsic::kDpdxCoarse || i == Intrinsic::kDpdyCoarse || - i == Intrinsic::kFwidthCoarse; +bool IsCoarseDerivative(IntrinsicType i) { + return i == IntrinsicType::kDpdxCoarse || i == IntrinsicType::kDpdyCoarse || + i == IntrinsicType::kFwidthCoarse; } -bool IsFineDerivative(Intrinsic i) { - return i == Intrinsic::kDpdxFine || i == Intrinsic::kDpdyFine || - i == Intrinsic::kFwidthFine; +bool IsFineDerivative(IntrinsicType i) { + return i == IntrinsicType::kDpdxFine || i == IntrinsicType::kDpdyFine || + i == IntrinsicType::kFwidthFine; } -bool IsDerivative(Intrinsic i) { - return i == Intrinsic::kDpdx || i == Intrinsic::kDpdy || - i == Intrinsic::kFwidth || IsCoarseDerivative(i) || +bool IsDerivative(IntrinsicType i) { + return i == IntrinsicType::kDpdx || i == IntrinsicType::kDpdy || + i == IntrinsicType::kFwidth || IsCoarseDerivative(i) || IsFineDerivative(i); } -bool IsFloatClassificationIntrinsic(Intrinsic i) { - return i == Intrinsic::kIsFinite || i == Intrinsic::kIsInf || - i == Intrinsic::kIsNan || i == Intrinsic::kIsNormal; +bool IsFloatClassificationIntrinsic(IntrinsicType i) { + return i == IntrinsicType::kIsFinite || i == IntrinsicType::kIsInf || + i == IntrinsicType::kIsNan || i == IntrinsicType::kIsNormal; } -bool IsTextureIntrinsic(Intrinsic i) { - return IsImageQueryIntrinsic(i) || i == Intrinsic::kTextureLoad || - i == Intrinsic::kTextureSample || - i == Intrinsic::kTextureSampleLevel || - i == Intrinsic::kTextureSampleBias || - i == Intrinsic::kTextureSampleCompare || - i == Intrinsic::kTextureSampleGrad || i == Intrinsic::kTextureStore; +bool IsTextureIntrinsic(IntrinsicType i) { + return IsImageQueryIntrinsic(i) || i == IntrinsicType::kTextureLoad || + i == IntrinsicType::kTextureSample || + i == IntrinsicType::kTextureSampleLevel || + i == IntrinsicType::kTextureSampleBias || + i == IntrinsicType::kTextureSampleCompare || + i == IntrinsicType::kTextureSampleGrad || + i == IntrinsicType::kTextureStore; } -bool IsImageQueryIntrinsic(Intrinsic i) { - return i == semantic::Intrinsic::kTextureDimensions || - i == Intrinsic::kTextureNumLayers || - i == Intrinsic::kTextureNumLevels || - i == Intrinsic::kTextureNumSamples; +bool IsImageQueryIntrinsic(IntrinsicType i) { + return i == semantic::IntrinsicType::kTextureDimensions || + i == IntrinsicType::kTextureNumLayers || + i == IntrinsicType::kTextureNumLevels || + i == IntrinsicType::kTextureNumSamples; } -bool IsDataPackingIntrinsic(Intrinsic i) { - return i == Intrinsic::kPack4x8Snorm || i == Intrinsic::kPack4x8Unorm || - i == Intrinsic::kPack2x16Snorm || i == Intrinsic::kPack2x16Unorm || - i == Intrinsic::kPack2x16Float; +bool IsDataPackingIntrinsic(IntrinsicType i) { + return i == IntrinsicType::kPack4x8Snorm || + i == IntrinsicType::kPack4x8Unorm || + i == IntrinsicType::kPack2x16Snorm || + i == IntrinsicType::kPack2x16Unorm || + i == IntrinsicType::kPack2x16Float; } } // namespace intrinsic diff --git a/src/semantic/intrinsic.h b/src/semantic/intrinsic.h index fd5a544797..f006f31ef2 100644 --- a/src/semantic/intrinsic.h +++ b/src/semantic/intrinsic.h @@ -20,7 +20,7 @@ namespace tint { namespace semantic { -enum class Intrinsic { +enum class IntrinsicType { kNone = -1, kAbs, @@ -104,48 +104,48 @@ enum class Intrinsic { /// Emits the name of the intrinsic function. The spelling, /// including case, matches the name in the WGSL spec. -std::ostream& operator<<(std::ostream& out, Intrinsic i); +std::ostream& operator<<(std::ostream& out, IntrinsicType i); namespace intrinsic { /// Determines if the given `i` is a coarse derivative /// @param i the intrinsic /// @returns true if the given derivative is coarse. -bool IsCoarseDerivative(Intrinsic i); +bool IsCoarseDerivative(IntrinsicType i); /// Determines if the given `i` is a fine derivative /// @param i the intrinsic /// @returns true if the given derivative is fine. -bool IsFineDerivative(Intrinsic i); +bool IsFineDerivative(IntrinsicType i); /// Determine if the given `i` is a derivative intrinsic /// @param i the intrinsic /// @returns true if the given `i` is a derivative intrinsic -bool IsDerivative(Intrinsic i); +bool IsDerivative(IntrinsicType i); /// Determines if the given `i` is a float classification intrinsic /// @param i the intrinsic /// @returns true if the given `i` is a float intrinsic -bool IsFloatClassificationIntrinsic(Intrinsic i); +bool IsFloatClassificationIntrinsic(IntrinsicType i); /// Determines if the given `i` is a texture operation intrinsic /// @param i the intrinsic /// @returns true if the given `i` is a texture operation intrinsic -bool IsTextureIntrinsic(Intrinsic i); +bool IsTextureIntrinsic(IntrinsicType i); /// Determines if the given `i` is a image query intrinsic /// @param i the intrinsic /// @returns true if the given `i` is a image query intrinsic -bool IsImageQueryIntrinsic(Intrinsic i); +bool IsImageQueryIntrinsic(IntrinsicType i); /// Determines if the given `i` is a data packing intrinsic /// @param i the intrinsic /// @returns true if the given `i` is a data packing intrinsic -bool IsDataPackingIntrinsic(Intrinsic i); +bool IsDataPackingIntrinsic(IntrinsicType i); /// @returns the name of the intrinsic function. The spelling, including case, /// matches the name in the WGSL spec. -const char* str(Intrinsic i); +const char* str(IntrinsicType i); } // namespace intrinsic } // namespace semantic diff --git a/src/semantic/sem_call.cc b/src/semantic/sem_call.cc index 72be60a57c..777ad51630 100644 --- a/src/semantic/sem_call.cc +++ b/src/semantic/sem_call.cc @@ -25,13 +25,13 @@ Call::Call(type::Type* return_type) : Base(return_type) {} Call::~Call() = default; -IntrinsicCall::IntrinsicCall(type::Type* return_type, Intrinsic intrinsic) +IntrinsicCall::IntrinsicCall(type::Type* return_type, IntrinsicType intrinsic) : Base(return_type), intrinsic_(intrinsic) {} IntrinsicCall::~IntrinsicCall() = default; TextureIntrinsicCall::TextureIntrinsicCall(type::Type* return_type, - Intrinsic intrinsic, + IntrinsicType intrinsic, const Parameters& params) : Base(return_type, intrinsic), params_(params) {} diff --git a/src/type_determiner.cc b/src/type_determiner.cc index 2e72f4af5b..6b47ffa587 100644 --- a/src/type_determiner.cc +++ b/src/type_determiner.cc @@ -65,6 +65,11 @@ #include "src/type/void_type.h" namespace tint { +namespace { + +using IntrinsicType = tint::semantic::IntrinsicType; + +} // namespace TypeDeterminer::TypeDeterminer(ProgramBuilder* builder) : builder_(builder) {} @@ -415,7 +420,7 @@ bool TypeDeterminer::DetermineCall(ast::CallExpression* call) { auto name = builder_->Symbols().NameFor(ident->symbol()); auto intrinsic = MatchIntrinsic(name); - if (intrinsic != semantic::Intrinsic::kNone) { + if (intrinsic != IntrinsicType::kNone) { if (!DetermineIntrinsicCall(call, intrinsic)) { return false; } @@ -464,7 +469,7 @@ enum class IntrinsicDataType { }; struct IntrinsicData { - semantic::Intrinsic intrinsic; + IntrinsicType intrinsic; IntrinsicDataType result_type; uint8_t result_vector_width; uint8_t param_for_result_type; @@ -473,74 +478,68 @@ struct IntrinsicData { // Note, this isn't all the intrinsics. Some are handled specially before // we get to the generic code. See the DetermineIntrinsic code below. constexpr const IntrinsicData kIntrinsicData[] = { - {semantic::Intrinsic::kAbs, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kAcos, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kAll, IntrinsicDataType::kBool, 1, 0}, - {semantic::Intrinsic::kAny, IntrinsicDataType::kBool, 1, 0}, - {semantic::Intrinsic::kArrayLength, IntrinsicDataType::kUnsignedInteger, 1, - 0}, - {semantic::Intrinsic::kAsin, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kAtan, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kAtan2, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kCeil, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kClamp, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kCos, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kCosh, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kCountOneBits, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kCross, IntrinsicDataType::kFloat, 3, 0}, - {semantic::Intrinsic::kDeterminant, IntrinsicDataType::kFloat, 1, 0}, - {semantic::Intrinsic::kDistance, IntrinsicDataType::kFloat, 1, 0}, - {semantic::Intrinsic::kDpdx, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kDpdxCoarse, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kDpdxFine, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kDpdy, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kDpdyCoarse, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kDpdyFine, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kDot, IntrinsicDataType::kFloat, 1, 0}, - {semantic::Intrinsic::kExp, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kExp2, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFaceForward, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFloor, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFwidth, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFwidthCoarse, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFwidthFine, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFma, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFract, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kFrexp, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kInverseSqrt, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kLdexp, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kLength, IntrinsicDataType::kFloat, 1, 0}, - {semantic::Intrinsic::kLog, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kLog2, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kMax, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kMin, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kMix, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kModf, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kNormalize, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kPack4x8Snorm, IntrinsicDataType::kUnsignedInteger, 1, - 0}, - {semantic::Intrinsic::kPack4x8Unorm, IntrinsicDataType::kUnsignedInteger, 1, - 0}, - {semantic::Intrinsic::kPack2x16Snorm, IntrinsicDataType::kUnsignedInteger, - 1, 0}, - {semantic::Intrinsic::kPack2x16Unorm, IntrinsicDataType::kUnsignedInteger, - 1, 0}, - {semantic::Intrinsic::kPack2x16Float, IntrinsicDataType::kUnsignedInteger, - 1, 0}, - {semantic::Intrinsic::kPow, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kReflect, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kReverseBits, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kRound, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kSelect, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kSign, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kSin, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kSinh, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kSmoothStep, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kSqrt, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kStep, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kTan, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kTanh, IntrinsicDataType::kDependent, 0, 0}, - {semantic::Intrinsic::kTrunc, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kAbs, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kAcos, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kAll, IntrinsicDataType::kBool, 1, 0}, + {IntrinsicType::kAny, IntrinsicDataType::kBool, 1, 0}, + {IntrinsicType::kArrayLength, IntrinsicDataType::kUnsignedInteger, 1, 0}, + {IntrinsicType::kAsin, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kAtan, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kAtan2, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kCeil, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kClamp, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kCos, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kCosh, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kCountOneBits, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kCross, IntrinsicDataType::kFloat, 3, 0}, + {IntrinsicType::kDeterminant, IntrinsicDataType::kFloat, 1, 0}, + {IntrinsicType::kDistance, IntrinsicDataType::kFloat, 1, 0}, + {IntrinsicType::kDpdx, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kDpdxCoarse, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kDpdxFine, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kDpdy, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kDpdyCoarse, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kDpdyFine, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kDot, IntrinsicDataType::kFloat, 1, 0}, + {IntrinsicType::kExp, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kExp2, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFaceForward, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFloor, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFwidth, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFwidthCoarse, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFwidthFine, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFma, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFract, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kFrexp, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kInverseSqrt, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kLdexp, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kLength, IntrinsicDataType::kFloat, 1, 0}, + {IntrinsicType::kLog, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kLog2, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kMax, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kMin, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kMix, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kModf, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kNormalize, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kPack4x8Snorm, IntrinsicDataType::kUnsignedInteger, 1, 0}, + {IntrinsicType::kPack4x8Unorm, IntrinsicDataType::kUnsignedInteger, 1, 0}, + {IntrinsicType::kPack2x16Snorm, IntrinsicDataType::kUnsignedInteger, 1, 0}, + {IntrinsicType::kPack2x16Unorm, IntrinsicDataType::kUnsignedInteger, 1, 0}, + {IntrinsicType::kPack2x16Float, IntrinsicDataType::kUnsignedInteger, 1, 0}, + {IntrinsicType::kPow, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kReflect, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kReverseBits, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kRound, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kSelect, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kSign, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kSin, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kSinh, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kSmoothStep, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kSqrt, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kStep, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kTan, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kTanh, IntrinsicDataType::kDependent, 0, 0}, + {IntrinsicType::kTrunc, IntrinsicDataType::kDependent, 0, 0}, }; constexpr const uint32_t kIntrinsicDataCount = @@ -549,7 +548,7 @@ constexpr const uint32_t kIntrinsicDataCount = } // namespace bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, - semantic::Intrinsic intrinsic) { + IntrinsicType intrinsic) { auto create_sem = [&](type::Type* result) { auto* sem = builder_->create(result, intrinsic); builder_->Sem().Add(call, sem); @@ -586,18 +585,18 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, bool is_array = type::IsTextureArray(texture->dim()); bool is_multisampled = texture->Is(); switch (intrinsic) { - case semantic::Intrinsic::kTextureDimensions: + case IntrinsicType::kTextureDimensions: param.idx.texture = param.count++; if (call->params().size() > param.count) { param.idx.level = param.count++; } break; - case semantic::Intrinsic::kTextureNumLayers: - case semantic::Intrinsic::kTextureNumLevels: - case semantic::Intrinsic::kTextureNumSamples: + case IntrinsicType::kTextureNumLayers: + case IntrinsicType::kTextureNumLevels: + case IntrinsicType::kTextureNumSamples: param.idx.texture = param.count++; break; - case semantic::Intrinsic::kTextureLoad: + case IntrinsicType::kTextureLoad: param.idx.texture = param.count++; param.idx.coords = param.count++; if (is_array) { @@ -611,7 +610,7 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, } } break; - case semantic::Intrinsic::kTextureSample: + case IntrinsicType::kTextureSample: param.idx.texture = param.count++; param.idx.sampler = param.count++; param.idx.coords = param.count++; @@ -622,7 +621,7 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, param.idx.offset = param.count++; } break; - case semantic::Intrinsic::kTextureSampleBias: + case IntrinsicType::kTextureSampleBias: param.idx.texture = param.count++; param.idx.sampler = param.count++; param.idx.coords = param.count++; @@ -634,7 +633,7 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, param.idx.offset = param.count++; } break; - case semantic::Intrinsic::kTextureSampleLevel: + case IntrinsicType::kTextureSampleLevel: param.idx.texture = param.count++; param.idx.sampler = param.count++; param.idx.coords = param.count++; @@ -646,7 +645,7 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, param.idx.offset = param.count++; } break; - case semantic::Intrinsic::kTextureSampleCompare: + case IntrinsicType::kTextureSampleCompare: param.idx.texture = param.count++; param.idx.sampler = param.count++; param.idx.coords = param.count++; @@ -658,7 +657,7 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, param.idx.offset = param.count++; } break; - case semantic::Intrinsic::kTextureSampleGrad: + case IntrinsicType::kTextureSampleGrad: param.idx.texture = param.count++; param.idx.sampler = param.count++; param.idx.coords = param.count++; @@ -671,7 +670,7 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, param.idx.offset = param.count++; } break; - case semantic::Intrinsic::kTextureStore: + case IntrinsicType::kTextureStore: param.idx.texture = param.count++; param.idx.coords = param.count++; if (is_array) { @@ -697,7 +696,7 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, // Set the function return type type::Type* return_type = nullptr; switch (intrinsic) { - case semantic::Intrinsic::kTextureDimensions: { + case IntrinsicType::kTextureDimensions: { auto* i32 = builder_->create(); switch (texture->dim()) { default: @@ -719,12 +718,12 @@ bool TypeDeterminer::DetermineIntrinsicCall(ast::CallExpression* call, } break; } - case semantic::Intrinsic::kTextureNumLayers: - case semantic::Intrinsic::kTextureNumLevels: - case semantic::Intrinsic::kTextureNumSamples: + case IntrinsicType::kTextureNumLayers: + case IntrinsicType::kTextureNumLevels: + case IntrinsicType::kTextureNumSamples: return_type = builder_->create(); break; - case semantic::Intrinsic::kTextureStore: + case IntrinsicType::kTextureStore: return_type = builder_->create(); break; default: { @@ -848,7 +847,7 @@ bool TypeDeterminer::DetermineIdentifier(ast::IdentifierExpression* expr) { } std::string name = builder_->Symbols().NameFor(symbol); - if (MatchIntrinsic(name) != semantic::Intrinsic::kNone) { + if (MatchIntrinsic(name) != IntrinsicType::kNone) { // Identifier is to an intrinsic function, which has no type (currently). return true; } @@ -858,163 +857,163 @@ bool TypeDeterminer::DetermineIdentifier(ast::IdentifierExpression* expr) { return false; } -semantic::Intrinsic TypeDeterminer::MatchIntrinsic(const std::string& name) { +IntrinsicType TypeDeterminer::MatchIntrinsic(const std::string& name) { if (name == "abs") { - return semantic::Intrinsic::kAbs; + return IntrinsicType::kAbs; } else if (name == "acos") { - return semantic::Intrinsic::kAcos; + return IntrinsicType::kAcos; } else if (name == "all") { - return semantic::Intrinsic::kAll; + return IntrinsicType::kAll; } else if (name == "any") { - return semantic::Intrinsic::kAny; + return IntrinsicType::kAny; } else if (name == "arrayLength") { - return semantic::Intrinsic::kArrayLength; + return IntrinsicType::kArrayLength; } else if (name == "asin") { - return semantic::Intrinsic::kAsin; + return IntrinsicType::kAsin; } else if (name == "atan") { - return semantic::Intrinsic::kAtan; + return IntrinsicType::kAtan; } else if (name == "atan2") { - return semantic::Intrinsic::kAtan2; + return IntrinsicType::kAtan2; } else if (name == "ceil") { - return semantic::Intrinsic::kCeil; + return IntrinsicType::kCeil; } else if (name == "clamp") { - return semantic::Intrinsic::kClamp; + return IntrinsicType::kClamp; } else if (name == "cos") { - return semantic::Intrinsic::kCos; + return IntrinsicType::kCos; } else if (name == "cosh") { - return semantic::Intrinsic::kCosh; + return IntrinsicType::kCosh; } else if (name == "countOneBits") { - return semantic::Intrinsic::kCountOneBits; + return IntrinsicType::kCountOneBits; } else if (name == "cross") { - return semantic::Intrinsic::kCross; + return IntrinsicType::kCross; } else if (name == "determinant") { - return semantic::Intrinsic::kDeterminant; + return IntrinsicType::kDeterminant; } else if (name == "distance") { - return semantic::Intrinsic::kDistance; + return IntrinsicType::kDistance; } else if (name == "dot") { - return semantic::Intrinsic::kDot; + return IntrinsicType::kDot; } else if (name == "dpdx") { - return semantic::Intrinsic::kDpdx; + return IntrinsicType::kDpdx; } else if (name == "dpdxCoarse") { - return semantic::Intrinsic::kDpdxCoarse; + return IntrinsicType::kDpdxCoarse; } else if (name == "dpdxFine") { - return semantic::Intrinsic::kDpdxFine; + return IntrinsicType::kDpdxFine; } else if (name == "dpdy") { - return semantic::Intrinsic::kDpdy; + return IntrinsicType::kDpdy; } else if (name == "dpdyCoarse") { - return semantic::Intrinsic::kDpdyCoarse; + return IntrinsicType::kDpdyCoarse; } else if (name == "dpdyFine") { - return semantic::Intrinsic::kDpdyFine; + return IntrinsicType::kDpdyFine; } else if (name == "exp") { - return semantic::Intrinsic::kExp; + return IntrinsicType::kExp; } else if (name == "exp2") { - return semantic::Intrinsic::kExp2; + return IntrinsicType::kExp2; } else if (name == "faceForward") { - return semantic::Intrinsic::kFaceForward; + return IntrinsicType::kFaceForward; } else if (name == "floor") { - return semantic::Intrinsic::kFloor; + return IntrinsicType::kFloor; } else if (name == "fma") { - return semantic::Intrinsic::kFma; + return IntrinsicType::kFma; } else if (name == "fract") { - return semantic::Intrinsic::kFract; + return IntrinsicType::kFract; } else if (name == "frexp") { - return semantic::Intrinsic::kFrexp; + return IntrinsicType::kFrexp; } else if (name == "fwidth") { - return semantic::Intrinsic::kFwidth; + return IntrinsicType::kFwidth; } else if (name == "fwidthCoarse") { - return semantic::Intrinsic::kFwidthCoarse; + return IntrinsicType::kFwidthCoarse; } else if (name == "fwidthFine") { - return semantic::Intrinsic::kFwidthFine; + return IntrinsicType::kFwidthFine; } else if (name == "inverseSqrt") { - return semantic::Intrinsic::kInverseSqrt; + return IntrinsicType::kInverseSqrt; } else if (name == "isFinite") { - return semantic::Intrinsic::kIsFinite; + return IntrinsicType::kIsFinite; } else if (name == "isInf") { - return semantic::Intrinsic::kIsInf; + return IntrinsicType::kIsInf; } else if (name == "isNan") { - return semantic::Intrinsic::kIsNan; + return IntrinsicType::kIsNan; } else if (name == "isNormal") { - return semantic::Intrinsic::kIsNormal; + return IntrinsicType::kIsNormal; } else if (name == "ldexp") { - return semantic::Intrinsic::kLdexp; + return IntrinsicType::kLdexp; } else if (name == "length") { - return semantic::Intrinsic::kLength; + return IntrinsicType::kLength; } else if (name == "log") { - return semantic::Intrinsic::kLog; + return IntrinsicType::kLog; } else if (name == "log2") { - return semantic::Intrinsic::kLog2; + return IntrinsicType::kLog2; } else if (name == "max") { - return semantic::Intrinsic::kMax; + return IntrinsicType::kMax; } else if (name == "min") { - return semantic::Intrinsic::kMin; + return IntrinsicType::kMin; } else if (name == "mix") { - return semantic::Intrinsic::kMix; + return IntrinsicType::kMix; } else if (name == "modf") { - return semantic::Intrinsic::kModf; + return IntrinsicType::kModf; } else if (name == "normalize") { - return semantic::Intrinsic::kNormalize; + return IntrinsicType::kNormalize; } else if (name == "pack4x8snorm") { - return semantic::Intrinsic::kPack4x8Snorm; + return IntrinsicType::kPack4x8Snorm; } else if (name == "pack4x8unorm") { - return semantic::Intrinsic::kPack4x8Unorm; + return IntrinsicType::kPack4x8Unorm; } else if (name == "pack2x16snorm") { - return semantic::Intrinsic::kPack2x16Snorm; + return IntrinsicType::kPack2x16Snorm; } else if (name == "pack2x16unorm") { - return semantic::Intrinsic::kPack2x16Unorm; + return IntrinsicType::kPack2x16Unorm; } else if (name == "pack2x16float") { - return semantic::Intrinsic::kPack2x16Float; + return IntrinsicType::kPack2x16Float; } else if (name == "pow") { - return semantic::Intrinsic::kPow; + return IntrinsicType::kPow; } else if (name == "reflect") { - return semantic::Intrinsic::kReflect; + return IntrinsicType::kReflect; } else if (name == "reverseBits") { - return semantic::Intrinsic::kReverseBits; + return IntrinsicType::kReverseBits; } else if (name == "round") { - return semantic::Intrinsic::kRound; + return IntrinsicType::kRound; } else if (name == "select") { - return semantic::Intrinsic::kSelect; + return IntrinsicType::kSelect; } else if (name == "sign") { - return semantic::Intrinsic::kSign; + return IntrinsicType::kSign; } else if (name == "sin") { - return semantic::Intrinsic::kSin; + return IntrinsicType::kSin; } else if (name == "sinh") { - return semantic::Intrinsic::kSinh; + return IntrinsicType::kSinh; } else if (name == "smoothStep") { - return semantic::Intrinsic::kSmoothStep; + return IntrinsicType::kSmoothStep; } else if (name == "sqrt") { - return semantic::Intrinsic::kSqrt; + return IntrinsicType::kSqrt; } else if (name == "step") { - return semantic::Intrinsic::kStep; + return IntrinsicType::kStep; } else if (name == "tan") { - return semantic::Intrinsic::kTan; + return IntrinsicType::kTan; } else if (name == "tanh") { - return semantic::Intrinsic::kTanh; + return IntrinsicType::kTanh; } else if (name == "textureDimensions") { - return semantic::Intrinsic::kTextureDimensions; + return IntrinsicType::kTextureDimensions; } else if (name == "textureNumLayers") { - return semantic::Intrinsic::kTextureNumLayers; + return IntrinsicType::kTextureNumLayers; } else if (name == "textureNumLevels") { - return semantic::Intrinsic::kTextureNumLevels; + return IntrinsicType::kTextureNumLevels; } else if (name == "textureNumSamples") { - return semantic::Intrinsic::kTextureNumSamples; + return IntrinsicType::kTextureNumSamples; } else if (name == "textureLoad") { - return semantic::Intrinsic::kTextureLoad; + return IntrinsicType::kTextureLoad; } else if (name == "textureStore") { - return semantic::Intrinsic::kTextureStore; + return IntrinsicType::kTextureStore; } else if (name == "textureSample") { - return semantic::Intrinsic::kTextureSample; + return IntrinsicType::kTextureSample; } else if (name == "textureSampleBias") { - return semantic::Intrinsic::kTextureSampleBias; + return IntrinsicType::kTextureSampleBias; } else if (name == "textureSampleCompare") { - return semantic::Intrinsic::kTextureSampleCompare; + return IntrinsicType::kTextureSampleCompare; } else if (name == "textureSampleGrad") { - return semantic::Intrinsic::kTextureSampleGrad; + return IntrinsicType::kTextureSampleGrad; } else if (name == "textureSampleLevel") { - return semantic::Intrinsic::kTextureSampleLevel; + return IntrinsicType::kTextureSampleLevel; } else if (name == "trunc") { - return semantic::Intrinsic::kTrunc; + return IntrinsicType::kTrunc; } - return semantic::Intrinsic::kNone; + return IntrinsicType::kNone; } bool TypeDeterminer::DetermineMemberAccessor( diff --git a/src/type_determiner.h b/src/type_determiner.h index 916a6b28e9..c8dc34d297 100644 --- a/src/type_determiner.h +++ b/src/type_determiner.h @@ -67,9 +67,9 @@ class TypeDeterminer { bool Determine(); /// @param name the function name to try and match as an intrinsic. - /// @return the semantic::Intrinsic for the given name. If `name` does not - /// match an intrinsic, returns semantic::Intrinsic::kNone - static semantic::Intrinsic MatchIntrinsic(const std::string& name); + /// @return the semantic::IntrinsicType for the given name. If `name` does not + /// match an intrinsic, returns semantic::IntrinsicType::kNone + static semantic::IntrinsicType MatchIntrinsic(const std::string& name); private: template @@ -177,7 +177,7 @@ class TypeDeterminer { bool DetermineConstructor(ast::ConstructorExpression* expr); bool DetermineIdentifier(ast::IdentifierExpression* expr); bool DetermineIntrinsicCall(ast::CallExpression* call, - semantic::Intrinsic intrinsic); + semantic::IntrinsicType intrinsic); bool DetermineMemberAccessor(ast::MemberAccessorExpression* expr); bool DetermineUnaryOp(ast::UnaryOpExpression* expr); diff --git a/src/type_determiner_test.cc b/src/type_determiner_test.cc index c6822c1797..9013905d53 100644 --- a/src/type_determiner_test.cc +++ b/src/type_determiner_test.cc @@ -75,6 +75,8 @@ namespace tint { namespace { +using IntrinsicType = semantic::IntrinsicType; + class FakeStmt : public ast::Statement { public: explicit FakeStmt(Source source) : ast::Statement(source) {} @@ -1618,7 +1620,7 @@ TEST_F(TypeDeterminerTest, StorageClass_NonFunctionClassError) { struct IntrinsicData { const char* name; - semantic::Intrinsic intrinsic; + IntrinsicType intrinsic; }; inline std::ostream& operator<<(std::ostream& out, IntrinsicData data) { out << data.name; @@ -1634,89 +1636,82 @@ INSTANTIATE_TEST_SUITE_P( TypeDeterminerTest, IntrinsicDataTest, testing::Values( - IntrinsicData{"abs", semantic::Intrinsic::kAbs}, - IntrinsicData{"acos", semantic::Intrinsic::kAcos}, - IntrinsicData{"all", semantic::Intrinsic::kAll}, - IntrinsicData{"any", semantic::Intrinsic::kAny}, - IntrinsicData{"arrayLength", semantic::Intrinsic::kArrayLength}, - IntrinsicData{"asin", semantic::Intrinsic::kAsin}, - IntrinsicData{"atan", semantic::Intrinsic::kAtan}, - IntrinsicData{"atan2", semantic::Intrinsic::kAtan2}, - IntrinsicData{"ceil", semantic::Intrinsic::kCeil}, - IntrinsicData{"clamp", semantic::Intrinsic::kClamp}, - IntrinsicData{"cos", semantic::Intrinsic::kCos}, - IntrinsicData{"cosh", semantic::Intrinsic::kCosh}, - IntrinsicData{"countOneBits", semantic::Intrinsic::kCountOneBits}, - IntrinsicData{"cross", semantic::Intrinsic::kCross}, - IntrinsicData{"determinant", semantic::Intrinsic::kDeterminant}, - IntrinsicData{"distance", semantic::Intrinsic::kDistance}, - IntrinsicData{"dot", semantic::Intrinsic::kDot}, - IntrinsicData{"dpdx", semantic::Intrinsic::kDpdx}, - IntrinsicData{"dpdxCoarse", semantic::Intrinsic::kDpdxCoarse}, - IntrinsicData{"dpdxFine", semantic::Intrinsic::kDpdxFine}, - IntrinsicData{"dpdy", semantic::Intrinsic::kDpdy}, - IntrinsicData{"dpdyCoarse", semantic::Intrinsic::kDpdyCoarse}, - IntrinsicData{"dpdyFine", semantic::Intrinsic::kDpdyFine}, - IntrinsicData{"exp", semantic::Intrinsic::kExp}, - IntrinsicData{"exp2", semantic::Intrinsic::kExp2}, - IntrinsicData{"faceForward", semantic::Intrinsic::kFaceForward}, - IntrinsicData{"floor", semantic::Intrinsic::kFloor}, - IntrinsicData{"fma", semantic::Intrinsic::kFma}, - IntrinsicData{"fract", semantic::Intrinsic::kFract}, - IntrinsicData{"frexp", semantic::Intrinsic::kFrexp}, - IntrinsicData{"fwidth", semantic::Intrinsic::kFwidth}, - IntrinsicData{"fwidthCoarse", semantic::Intrinsic::kFwidthCoarse}, - IntrinsicData{"fwidthFine", semantic::Intrinsic::kFwidthFine}, - IntrinsicData{"inverseSqrt", semantic::Intrinsic::kInverseSqrt}, - IntrinsicData{"isFinite", semantic::Intrinsic::kIsFinite}, - IntrinsicData{"isInf", semantic::Intrinsic::kIsInf}, - IntrinsicData{"isNan", semantic::Intrinsic::kIsNan}, - IntrinsicData{"isNormal", semantic::Intrinsic::kIsNormal}, - IntrinsicData{"ldexp", semantic::Intrinsic::kLdexp}, - IntrinsicData{"length", semantic::Intrinsic::kLength}, - IntrinsicData{"log", semantic::Intrinsic::kLog}, - IntrinsicData{"log2", semantic::Intrinsic::kLog2}, - IntrinsicData{"max", semantic::Intrinsic::kMax}, - IntrinsicData{"min", semantic::Intrinsic::kMin}, - IntrinsicData{"mix", semantic::Intrinsic::kMix}, - IntrinsicData{"modf", semantic::Intrinsic::kModf}, - IntrinsicData{"normalize", semantic::Intrinsic::kNormalize}, - IntrinsicData{"pow", semantic::Intrinsic::kPow}, - IntrinsicData{"reflect", semantic::Intrinsic::kReflect}, - IntrinsicData{"reverseBits", semantic::Intrinsic::kReverseBits}, - IntrinsicData{"round", semantic::Intrinsic::kRound}, - IntrinsicData{"select", semantic::Intrinsic::kSelect}, - IntrinsicData{"sign", semantic::Intrinsic::kSign}, - IntrinsicData{"sin", semantic::Intrinsic::kSin}, - IntrinsicData{"sinh", semantic::Intrinsic::kSinh}, - IntrinsicData{"smoothStep", semantic::Intrinsic::kSmoothStep}, - IntrinsicData{"sqrt", semantic::Intrinsic::kSqrt}, - IntrinsicData{"step", semantic::Intrinsic::kStep}, - IntrinsicData{"tan", semantic::Intrinsic::kTan}, - IntrinsicData{"tanh", semantic::Intrinsic::kTanh}, - IntrinsicData{"textureDimensions", - semantic::Intrinsic::kTextureDimensions}, - IntrinsicData{"textureLoad", semantic::Intrinsic::kTextureLoad}, - IntrinsicData{"textureNumLayers", - semantic::Intrinsic::kTextureNumLayers}, - IntrinsicData{"textureNumLevels", - semantic::Intrinsic::kTextureNumLevels}, - IntrinsicData{"textureNumSamples", - semantic::Intrinsic::kTextureNumSamples}, - IntrinsicData{"textureSample", semantic::Intrinsic::kTextureSample}, - IntrinsicData{"textureSampleBias", - semantic::Intrinsic::kTextureSampleBias}, + IntrinsicData{"abs", IntrinsicType::kAbs}, + IntrinsicData{"acos", IntrinsicType::kAcos}, + IntrinsicData{"all", IntrinsicType::kAll}, + IntrinsicData{"any", IntrinsicType::kAny}, + IntrinsicData{"arrayLength", IntrinsicType::kArrayLength}, + IntrinsicData{"asin", IntrinsicType::kAsin}, + IntrinsicData{"atan", IntrinsicType::kAtan}, + IntrinsicData{"atan2", IntrinsicType::kAtan2}, + IntrinsicData{"ceil", IntrinsicType::kCeil}, + IntrinsicData{"clamp", IntrinsicType::kClamp}, + IntrinsicData{"cos", IntrinsicType::kCos}, + IntrinsicData{"cosh", IntrinsicType::kCosh}, + IntrinsicData{"countOneBits", IntrinsicType::kCountOneBits}, + IntrinsicData{"cross", IntrinsicType::kCross}, + IntrinsicData{"determinant", IntrinsicType::kDeterminant}, + IntrinsicData{"distance", IntrinsicType::kDistance}, + IntrinsicData{"dot", IntrinsicType::kDot}, + IntrinsicData{"dpdx", IntrinsicType::kDpdx}, + IntrinsicData{"dpdxCoarse", IntrinsicType::kDpdxCoarse}, + IntrinsicData{"dpdxFine", IntrinsicType::kDpdxFine}, + IntrinsicData{"dpdy", IntrinsicType::kDpdy}, + IntrinsicData{"dpdyCoarse", IntrinsicType::kDpdyCoarse}, + IntrinsicData{"dpdyFine", IntrinsicType::kDpdyFine}, + IntrinsicData{"exp", IntrinsicType::kExp}, + IntrinsicData{"exp2", IntrinsicType::kExp2}, + IntrinsicData{"faceForward", IntrinsicType::kFaceForward}, + IntrinsicData{"floor", IntrinsicType::kFloor}, + IntrinsicData{"fma", IntrinsicType::kFma}, + IntrinsicData{"fract", IntrinsicType::kFract}, + IntrinsicData{"frexp", IntrinsicType::kFrexp}, + IntrinsicData{"fwidth", IntrinsicType::kFwidth}, + IntrinsicData{"fwidthCoarse", IntrinsicType::kFwidthCoarse}, + IntrinsicData{"fwidthFine", IntrinsicType::kFwidthFine}, + IntrinsicData{"inverseSqrt", IntrinsicType::kInverseSqrt}, + IntrinsicData{"isFinite", IntrinsicType::kIsFinite}, + IntrinsicData{"isInf", IntrinsicType::kIsInf}, + IntrinsicData{"isNan", IntrinsicType::kIsNan}, + IntrinsicData{"isNormal", IntrinsicType::kIsNormal}, + IntrinsicData{"ldexp", IntrinsicType::kLdexp}, + IntrinsicData{"length", IntrinsicType::kLength}, + IntrinsicData{"log", IntrinsicType::kLog}, + IntrinsicData{"log2", IntrinsicType::kLog2}, + IntrinsicData{"max", IntrinsicType::kMax}, + IntrinsicData{"min", IntrinsicType::kMin}, + IntrinsicData{"mix", IntrinsicType::kMix}, + IntrinsicData{"modf", IntrinsicType::kModf}, + IntrinsicData{"normalize", IntrinsicType::kNormalize}, + IntrinsicData{"pow", IntrinsicType::kPow}, + IntrinsicData{"reflect", IntrinsicType::kReflect}, + IntrinsicData{"reverseBits", IntrinsicType::kReverseBits}, + IntrinsicData{"round", IntrinsicType::kRound}, + IntrinsicData{"select", IntrinsicType::kSelect}, + IntrinsicData{"sign", IntrinsicType::kSign}, + IntrinsicData{"sin", IntrinsicType::kSin}, + IntrinsicData{"sinh", IntrinsicType::kSinh}, + IntrinsicData{"smoothStep", IntrinsicType::kSmoothStep}, + IntrinsicData{"sqrt", IntrinsicType::kSqrt}, + IntrinsicData{"step", IntrinsicType::kStep}, + IntrinsicData{"tan", IntrinsicType::kTan}, + IntrinsicData{"tanh", IntrinsicType::kTanh}, + IntrinsicData{"textureDimensions", IntrinsicType::kTextureDimensions}, + IntrinsicData{"textureLoad", IntrinsicType::kTextureLoad}, + IntrinsicData{"textureNumLayers", IntrinsicType::kTextureNumLayers}, + IntrinsicData{"textureNumLevels", IntrinsicType::kTextureNumLevels}, + IntrinsicData{"textureNumSamples", IntrinsicType::kTextureNumSamples}, + IntrinsicData{"textureSample", IntrinsicType::kTextureSample}, + IntrinsicData{"textureSampleBias", IntrinsicType::kTextureSampleBias}, IntrinsicData{"textureSampleCompare", - semantic::Intrinsic::kTextureSampleCompare}, - IntrinsicData{"textureSampleGrad", - semantic::Intrinsic::kTextureSampleGrad}, - IntrinsicData{"textureSampleLevel", - semantic::Intrinsic::kTextureSampleLevel}, - IntrinsicData{"trunc", semantic::Intrinsic::kTrunc})); + IntrinsicType::kTextureSampleCompare}, + IntrinsicData{"textureSampleGrad", IntrinsicType::kTextureSampleGrad}, + IntrinsicData{"textureSampleLevel", IntrinsicType::kTextureSampleLevel}, + IntrinsicData{"trunc", IntrinsicType::kTrunc})); TEST_F(TypeDeterminerTest, MatchIntrinsicNoMatch) { EXPECT_EQ(TypeDeterminer::MatchIntrinsic("not_intrinsic"), - semantic::Intrinsic::kNone); + IntrinsicType::kNone); } using ImportData_DataPackingTest = TypeDeterminerTestWithParam; @@ -1736,11 +1731,11 @@ INSTANTIATE_TEST_SUITE_P( TypeDeterminerTest, ImportData_DataPackingTest, testing::Values( - IntrinsicData{"pack4x8snorm", semantic::Intrinsic::kPack4x8Snorm}, - IntrinsicData{"pack4x8unorm", semantic::Intrinsic::kPack4x8Unorm}, - IntrinsicData{"pack2x16snorm", semantic::Intrinsic::kPack2x16Snorm}, - IntrinsicData{"pack2x16unorm", semantic::Intrinsic::kPack2x16Unorm}, - IntrinsicData{"pack2x16float", semantic::Intrinsic::kPack2x16Float})); + IntrinsicData{"pack4x8snorm", IntrinsicType::kPack4x8Snorm}, + IntrinsicData{"pack4x8unorm", IntrinsicType::kPack4x8Unorm}, + IntrinsicData{"pack2x16snorm", IntrinsicType::kPack2x16Snorm}, + IntrinsicData{"pack2x16unorm", IntrinsicType::kPack2x16Unorm}, + IntrinsicData{"pack2x16float", IntrinsicType::kPack2x16Float})); using ImportData_SingleParamTest = TypeDeterminerTestWithParam; TEST_P(ImportData_SingleParamTest, Scalar) { @@ -1786,29 +1781,28 @@ TEST_P(ImportData_SingleParamTest, Error_NoParams) { INSTANTIATE_TEST_SUITE_P( TypeDeterminerTest, ImportData_SingleParamTest, - testing::Values(IntrinsicData{"acos", semantic::Intrinsic::kAcos}, - IntrinsicData{"asin", semantic::Intrinsic::kAsin}, - IntrinsicData{"atan", semantic::Intrinsic::kAtan}, - IntrinsicData{"ceil", semantic::Intrinsic::kCeil}, - IntrinsicData{"cos", semantic::Intrinsic::kCos}, - IntrinsicData{"cosh", semantic::Intrinsic::kCosh}, - IntrinsicData{"exp", semantic::Intrinsic::kExp}, - IntrinsicData{"exp2", semantic::Intrinsic::kExp2}, - IntrinsicData{"floor", semantic::Intrinsic::kFloor}, - IntrinsicData{"fract", semantic::Intrinsic::kFract}, - IntrinsicData{"inverseSqrt", - semantic::Intrinsic::kInverseSqrt}, - IntrinsicData{"log", semantic::Intrinsic::kLog}, - IntrinsicData{"log2", semantic::Intrinsic::kLog2}, - IntrinsicData{"normalize", semantic::Intrinsic::kNormalize}, - IntrinsicData{"round", semantic::Intrinsic::kRound}, - IntrinsicData{"sign", semantic::Intrinsic::kSign}, - IntrinsicData{"sin", semantic::Intrinsic::kSin}, - IntrinsicData{"sinh", semantic::Intrinsic::kSinh}, - IntrinsicData{"sqrt", semantic::Intrinsic::kSqrt}, - IntrinsicData{"tan", semantic::Intrinsic::kTan}, - IntrinsicData{"tanh", semantic::Intrinsic::kTanh}, - IntrinsicData{"trunc", semantic::Intrinsic::kTrunc})); + testing::Values(IntrinsicData{"acos", IntrinsicType::kAcos}, + IntrinsicData{"asin", IntrinsicType::kAsin}, + IntrinsicData{"atan", IntrinsicType::kAtan}, + IntrinsicData{"ceil", IntrinsicType::kCeil}, + IntrinsicData{"cos", IntrinsicType::kCos}, + IntrinsicData{"cosh", IntrinsicType::kCosh}, + IntrinsicData{"exp", IntrinsicType::kExp}, + IntrinsicData{"exp2", IntrinsicType::kExp2}, + IntrinsicData{"floor", IntrinsicType::kFloor}, + IntrinsicData{"fract", IntrinsicType::kFract}, + IntrinsicData{"inverseSqrt", IntrinsicType::kInverseSqrt}, + IntrinsicData{"log", IntrinsicType::kLog}, + IntrinsicData{"log2", IntrinsicType::kLog2}, + IntrinsicData{"normalize", IntrinsicType::kNormalize}, + IntrinsicData{"round", IntrinsicType::kRound}, + IntrinsicData{"sign", IntrinsicType::kSign}, + IntrinsicData{"sin", IntrinsicType::kSin}, + IntrinsicData{"sinh", IntrinsicType::kSinh}, + IntrinsicData{"sqrt", IntrinsicType::kSqrt}, + IntrinsicData{"tan", IntrinsicType::kTan}, + IntrinsicData{"tanh", IntrinsicType::kTanh}, + IntrinsicData{"trunc", IntrinsicType::kTrunc})); using ImportData_SingleParam_FloatOrInt_Test = TypeDeterminerTestWithParam; @@ -1919,8 +1913,8 @@ TEST_P(ImportData_SingleParam_FloatOrInt_Test, Error_NoParams) { INSTANTIATE_TEST_SUITE_P(TypeDeterminerTest, ImportData_SingleParam_FloatOrInt_Test, - testing::Values(IntrinsicData{ - "abs", semantic::Intrinsic::kAbs})); + testing::Values(IntrinsicData{"abs", + IntrinsicType::kAbs})); TEST_F(TypeDeterminerTest, ImportData_Length_Scalar) { auto* ident = Expr("length"); @@ -1992,10 +1986,10 @@ TEST_P(ImportData_TwoParamTest, Error_NoParams) { INSTANTIATE_TEST_SUITE_P( TypeDeterminerTest, ImportData_TwoParamTest, - testing::Values(IntrinsicData{"atan2", semantic::Intrinsic::kAtan2}, - IntrinsicData{"pow", semantic::Intrinsic::kPow}, - IntrinsicData{"step", semantic::Intrinsic::kStep}, - IntrinsicData{"reflect", semantic::Intrinsic::kReflect})); + testing::Values(IntrinsicData{"atan2", IntrinsicType::kAtan2}, + IntrinsicData{"pow", IntrinsicType::kPow}, + IntrinsicData{"step", IntrinsicType::kStep}, + IntrinsicData{"reflect", IntrinsicType::kReflect})); TEST_F(TypeDeterminerTest, ImportData_Distance_Scalar) { auto* ident = Expr("distance"); @@ -2093,11 +2087,10 @@ TEST_P(ImportData_ThreeParamTest, Error_NoParams) { INSTANTIATE_TEST_SUITE_P( TypeDeterminerTest, ImportData_ThreeParamTest, - testing::Values( - IntrinsicData{"mix", semantic::Intrinsic::kMix}, - IntrinsicData{"smoothStep", semantic::Intrinsic::kSmoothStep}, - IntrinsicData{"fma", semantic::Intrinsic::kFma}, - IntrinsicData{"faceForward", semantic::Intrinsic::kFaceForward})); + testing::Values(IntrinsicData{"mix", IntrinsicType::kMix}, + IntrinsicData{"smoothStep", IntrinsicType::kSmoothStep}, + IntrinsicData{"fma", IntrinsicType::kFma}, + IntrinsicData{"faceForward", IntrinsicType::kFaceForward})); using ImportData_ThreeParam_FloatOrInt_Test = TypeDeterminerTestWithParam; @@ -2200,8 +2193,8 @@ TEST_P(ImportData_ThreeParam_FloatOrInt_Test, Error_NoParams) { INSTANTIATE_TEST_SUITE_P(TypeDeterminerTest, ImportData_ThreeParam_FloatOrInt_Test, - testing::Values(IntrinsicData{ - "clamp", semantic::Intrinsic::kClamp})); + testing::Values(IntrinsicData{"clamp", + IntrinsicType::kClamp})); using ImportData_Int_SingleParamTest = TypeDeterminerTestWithParam; @@ -2248,9 +2241,8 @@ TEST_P(ImportData_Int_SingleParamTest, Error_NoParams) { INSTANTIATE_TEST_SUITE_P( TypeDeterminerTest, ImportData_Int_SingleParamTest, - testing::Values( - IntrinsicData{"countOneBits", semantic::Intrinsic::kCountOneBits}, - IntrinsicData{"reverseBits", semantic::Intrinsic::kReverseBits})); + testing::Values(IntrinsicData{"countOneBits", IntrinsicType::kCountOneBits}, + IntrinsicData{"reverseBits", IntrinsicType::kReverseBits})); using ImportData_FloatOrInt_TwoParamTest = TypeDeterminerTestWithParam; @@ -2351,8 +2343,8 @@ TEST_P(ImportData_FloatOrInt_TwoParamTest, Error_NoParams) { INSTANTIATE_TEST_SUITE_P( TypeDeterminerTest, ImportData_FloatOrInt_TwoParamTest, - testing::Values(IntrinsicData{"min", semantic::Intrinsic::kMin}, - IntrinsicData{"max", semantic::Intrinsic::kMax})); + testing::Values(IntrinsicData{"min", IntrinsicType::kMin}, + IntrinsicData{"max", IntrinsicType::kMax})); TEST_F(TypeDeterminerTest, ImportData_GLSL_Determinant) { Global("var", ast::StorageClass::kFunction, ty.mat3x3()); @@ -2385,8 +2377,7 @@ TEST_P(ImportData_Matrix_OneParam_Test, NoParams) { INSTANTIATE_TEST_SUITE_P(TypeDeterminerTest, ImportData_Matrix_OneParam_Test, testing::Values(IntrinsicData{ - "determinant", - semantic::Intrinsic::kDeterminant})); + "determinant", IntrinsicType::kDeterminant})); TEST_F(TypeDeterminerTest, Function_EntryPoints_StageDecoration) { // fn b() {} diff --git a/src/validator/validator_impl.cc b/src/validator/validator_impl.cc index bb568dedb3..16da3e3bd9 100644 --- a/src/validator/validator_impl.cc +++ b/src/validator/validator_impl.cc @@ -49,6 +49,8 @@ namespace tint { namespace { +using IntrinsicType = semantic::IntrinsicType; + enum class IntrinsicDataType { kMixed, kFloatOrIntScalarOrVector, @@ -63,7 +65,7 @@ enum class IntrinsicDataType { }; struct IntrinsicData { - semantic::Intrinsic intrinsic; + IntrinsicType intrinsic; uint32_t param_count; IntrinsicDataType data_type; uint32_t vector_size; @@ -73,122 +75,97 @@ struct IntrinsicData { // Note, this isn't all the intrinsics. Some are handled specially before // we get to the generic code. See the ValidateCallExpr code below. constexpr const IntrinsicData kIntrinsicData[] = { - {semantic::Intrinsic::kAbs, 1, IntrinsicDataType::kFloatOrIntScalarOrVector, - 0, true}, - {semantic::Intrinsic::kAcos, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kAbs, 1, IntrinsicDataType::kFloatOrIntScalarOrVector, 0, true}, - {semantic::Intrinsic::kAll, 1, IntrinsicDataType::kBoolVector, 0, false}, - {semantic::Intrinsic::kAny, 1, IntrinsicDataType::kBoolVector, 0, false}, - {semantic::Intrinsic::kArrayLength, 1, IntrinsicDataType::kMixed, 0, false}, - {semantic::Intrinsic::kAsin, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kAcos, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kAll, 1, IntrinsicDataType::kBoolVector, 0, false}, + {IntrinsicType::kAny, 1, IntrinsicDataType::kBoolVector, 0, false}, + {IntrinsicType::kArrayLength, 1, IntrinsicDataType::kMixed, 0, false}, + {IntrinsicType::kAsin, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kAtan, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kAtan2, 2, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kAtan, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kCeil, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kClamp, 3, IntrinsicDataType::kFloatOrIntScalarOrVector, 0, true}, - {semantic::Intrinsic::kAtan2, 2, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kCos, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kCosh, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kCountOneBits, 1, IntrinsicDataType::kIntScalarOrVector, 0, true}, - {semantic::Intrinsic::kCeil, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kClamp, 3, - IntrinsicDataType::kFloatOrIntScalarOrVector, 0, true}, - {semantic::Intrinsic::kCos, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kCosh, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kCountOneBits, 1, - IntrinsicDataType::kIntScalarOrVector, 0, true}, - {semantic::Intrinsic::kCross, 2, IntrinsicDataType::kFloatVector, 3, true}, - {semantic::Intrinsic::kDeterminant, 1, IntrinsicDataType::kMatrix, 0, + {IntrinsicType::kCross, 2, IntrinsicDataType::kFloatVector, 3, true}, + {IntrinsicType::kDeterminant, 1, IntrinsicDataType::kMatrix, 0, false}, + {IntrinsicType::kDistance, 2, IntrinsicDataType::kFloatScalarOrVector, 0, false}, - {semantic::Intrinsic::kDistance, 2, IntrinsicDataType::kFloatScalarOrVector, - 0, false}, - {semantic::Intrinsic::kDot, 2, IntrinsicDataType::kFloatVector, 0, false}, - {semantic::Intrinsic::kDpdx, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kDot, 2, IntrinsicDataType::kFloatVector, 0, false}, + {IntrinsicType::kDpdx, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kDpdxCoarse, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kDpdxCoarse, 1, - IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kDpdxFine, 1, IntrinsicDataType::kFloatScalarOrVector, + {IntrinsicType::kDpdxFine, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + true}, + {IntrinsicType::kDpdy, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kDpdyCoarse, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + true}, + {IntrinsicType::kDpdyFine, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + true}, + {IntrinsicType::kExp, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kExp2, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kFaceForward, 3, IntrinsicDataType::kFloatScalarOrVector, 0, + true}, + {IntrinsicType::kFloor, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + true}, + {IntrinsicType::kFma, 3, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kFract, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + true}, + {IntrinsicType::kFrexp, 2, IntrinsicDataType::kMixed, 0, false}, + {IntrinsicType::kFwidth, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + true}, + {IntrinsicType::kFwidthCoarse, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kDpdy, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kFwidthFine, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kDpdyCoarse, 1, - IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kDpdyFine, 1, IntrinsicDataType::kFloatScalarOrVector, - 0, true}, - {semantic::Intrinsic::kExp, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kInverseSqrt, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kExp2, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kLdexp, 2, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kFaceForward, 3, - IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kFloor, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kFma, 3, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kFract, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kFrexp, 2, IntrinsicDataType::kMixed, 0, false}, - {semantic::Intrinsic::kFwidth, 1, IntrinsicDataType::kFloatScalarOrVector, - 0, true}, - {semantic::Intrinsic::kFwidthCoarse, 1, - IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kFwidthFine, 1, - IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kInverseSqrt, 1, - IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kLdexp, 2, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kLength, 1, IntrinsicDataType::kFloatScalarOrVector, - 0, false}, - {semantic::Intrinsic::kLog, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kLog2, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kMax, 2, IntrinsicDataType::kFloatOrIntScalarOrVector, - 0, true}, - {semantic::Intrinsic::kMin, 2, IntrinsicDataType::kFloatOrIntScalarOrVector, - 0, true}, - {semantic::Intrinsic::kMix, 3, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kModf, 2, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kNormalize, 1, IntrinsicDataType::kFloatVector, 0, - true}, - {semantic::Intrinsic::kPack4x8Snorm, 1, IntrinsicDataType::kFloatVector, 4, + {IntrinsicType::kLength, 1, IntrinsicDataType::kFloatScalarOrVector, 0, false}, - {semantic::Intrinsic::kPack4x8Unorm, 1, IntrinsicDataType::kFloatVector, 4, + {IntrinsicType::kLog, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kLog2, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kMax, 2, IntrinsicDataType::kFloatOrIntScalarOrVector, 0, + true}, + {IntrinsicType::kMin, 2, IntrinsicDataType::kFloatOrIntScalarOrVector, 0, + true}, + {IntrinsicType::kMix, 3, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kModf, 2, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kNormalize, 1, IntrinsicDataType::kFloatVector, 0, true}, + {IntrinsicType::kPack4x8Snorm, 1, IntrinsicDataType::kFloatVector, 4, false}, - {semantic::Intrinsic::kPack2x16Snorm, 1, IntrinsicDataType::kFloatVector, 2, + {IntrinsicType::kPack4x8Unorm, 1, IntrinsicDataType::kFloatVector, 4, false}, - {semantic::Intrinsic::kPack2x16Unorm, 1, IntrinsicDataType::kFloatVector, 2, + {IntrinsicType::kPack2x16Snorm, 1, IntrinsicDataType::kFloatVector, 2, false}, - {semantic::Intrinsic::kPack2x16Float, 1, IntrinsicDataType::kFloatVector, 2, + {IntrinsicType::kPack2x16Unorm, 1, IntrinsicDataType::kFloatVector, 2, false}, - {semantic::Intrinsic::kPow, 2, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kPack2x16Float, 1, IntrinsicDataType::kFloatVector, 2, + false}, + {IntrinsicType::kPow, 2, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kReflect, 2, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kReflect, 2, IntrinsicDataType::kFloatScalarOrVector, - 0, true}, - {semantic::Intrinsic::kReverseBits, 1, - IntrinsicDataType::kIntScalarOrVector, 0, true}, - {semantic::Intrinsic::kRound, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kReverseBits, 1, IntrinsicDataType::kIntScalarOrVector, 0, true}, - {semantic::Intrinsic::kSelect, 3, IntrinsicDataType::kMixed, 0, false}, - {semantic::Intrinsic::kSign, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kRound, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kSin, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kSelect, 3, IntrinsicDataType::kMixed, 0, false}, + {IntrinsicType::kSign, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kSin, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kSinh, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kSmoothStep, 3, IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kSinh, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kSmoothStep, 3, - IntrinsicDataType::kFloatScalarOrVector, 0, true}, - {semantic::Intrinsic::kSqrt, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kStep, 2, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kTan, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kTanh, 1, IntrinsicDataType::kFloatScalarOrVector, 0, - true}, - {semantic::Intrinsic::kTrunc, 1, IntrinsicDataType::kFloatScalarOrVector, 0, + {IntrinsicType::kSqrt, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kStep, 2, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kTan, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kTanh, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, + {IntrinsicType::kTrunc, 1, IntrinsicDataType::kFloatScalarOrVector, 0, true}, }; @@ -731,7 +708,7 @@ bool ValidatorImpl::ValidateCallExpr(const ast::CallExpression* expr) { } } else { // Special cases. - if (data->intrinsic == semantic::Intrinsic::kFrexp) { + if (data->intrinsic == IntrinsicType::kFrexp) { auto* p0 = expr->params()[0]; auto* p1 = expr->params()[1]; auto* t0 = program_->TypeOf(p0)->UnwrapPtrIfNeeded(); @@ -773,7 +750,7 @@ bool ValidatorImpl::ValidateCallExpr(const ast::CallExpression* expr) { } } - if (data->intrinsic == semantic::Intrinsic::kSelect) { + if (data->intrinsic == IntrinsicType::kSelect) { auto* type = program_->TypeOf(expr); auto* t0 = program_->TypeOf(expr->params()[0])->UnwrapPtrIfNeeded(); auto* t1 = program_->TypeOf(expr->params()[1])->UnwrapPtrIfNeeded(); @@ -819,7 +796,7 @@ bool ValidatorImpl::ValidateCallExpr(const ast::CallExpression* expr) { } } - if (data->intrinsic == semantic::Intrinsic::kArrayLength) { + if (data->intrinsic == IntrinsicType::kArrayLength) { if (!program_->TypeOf(expr)->UnwrapPtrIfNeeded()->Is()) { add_error( expr->source(), @@ -840,15 +817,15 @@ bool ValidatorImpl::ValidateCallExpr(const ast::CallExpression* expr) { } // Result types don't match parameter types. - if (data->intrinsic == semantic::Intrinsic::kAll || - data->intrinsic == semantic::Intrinsic::kAny) { + if (data->intrinsic == IntrinsicType::kAll || + data->intrinsic == IntrinsicType::kAny) { if (!IsValidType(program_->TypeOf(expr), expr->source(), builtin, IntrinsicDataType::kBoolScalar, 0, this)) { return false; } } - if (data->intrinsic == semantic::Intrinsic::kDot) { + if (data->intrinsic == IntrinsicType::kDot) { if (!IsValidType(program_->TypeOf(expr), expr->source(), builtin, IntrinsicDataType::kFloatScalar, 0, this)) { return false; @@ -864,9 +841,9 @@ bool ValidatorImpl::ValidateCallExpr(const ast::CallExpression* expr) { } } - if (data->intrinsic == semantic::Intrinsic::kLength || - data->intrinsic == semantic::Intrinsic::kDistance || - data->intrinsic == semantic::Intrinsic::kDeterminant) { + if (data->intrinsic == IntrinsicType::kLength || + data->intrinsic == IntrinsicType::kDistance || + data->intrinsic == IntrinsicType::kDeterminant) { if (!IsValidType(program_->TypeOf(expr), expr->source(), builtin, IntrinsicDataType::kFloatScalar, 0, this)) { return false; @@ -874,7 +851,7 @@ bool ValidatorImpl::ValidateCallExpr(const ast::CallExpression* expr) { } // Must be a square matrix. - if (data->intrinsic == semantic::Intrinsic::kDeterminant) { + if (data->intrinsic == IntrinsicType::kDeterminant) { const auto* matrix = program_->TypeOf(expr->params()[0])->As(); if (matrix->rows() != matrix->columns()) { @@ -887,8 +864,8 @@ bool ValidatorImpl::ValidateCallExpr(const ast::CallExpression* expr) { } // Last parameter must be a pointer. - if (data->intrinsic == semantic::Intrinsic::kFrexp || - data->intrinsic == semantic::Intrinsic::kModf) { + if (data->intrinsic == IntrinsicType::kFrexp || + data->intrinsic == IntrinsicType::kModf) { auto* last_param = expr->params()[data->param_count - 1]; if (!program_->TypeOf(last_param)->Is()) { add_error(last_param->source(), "incorrect type for " + builtin + diff --git a/src/writer/hlsl/generator_impl.cc b/src/writer/hlsl/generator_impl.cc index 52dfa12715..2a4d304af3 100644 --- a/src/writer/hlsl/generator_impl.cc +++ b/src/writer/hlsl/generator_impl.cc @@ -548,10 +548,10 @@ bool GeneratorImpl::EmitCall(std::ostream& pre, } if (auto* sem = call_sem->As()) { const auto& params = expr->params(); - if (sem->intrinsic() == semantic::Intrinsic::kSelect) { + if (sem->intrinsic() == semantic::IntrinsicType::kSelect) { error_ = "select not supported in HLSL backend yet"; return false; - } else if (sem->intrinsic() == semantic::Intrinsic::kIsNormal) { + } else if (sem->intrinsic() == semantic::IntrinsicType::kIsNormal) { error_ = "is_normal not supported in HLSL backend yet"; return false; } else if (semantic::intrinsic::IsDataPackingIntrinsic(sem->intrinsic())) { @@ -650,21 +650,21 @@ bool GeneratorImpl::EmitDataPackingCall(std::ostream& pre, uint32_t dims = 2; bool is_signed = false; uint32_t scale = 65535; - if (ident->intrinsic() == semantic::Intrinsic::kPack4x8Snorm || - ident->intrinsic() == semantic::Intrinsic::kPack4x8Unorm) { + if (ident->intrinsic() == semantic::IntrinsicType::kPack4x8Snorm || + ident->intrinsic() == semantic::IntrinsicType::kPack4x8Unorm) { dims = 4; scale = 255; } - if (ident->intrinsic() == semantic::Intrinsic::kPack4x8Snorm || - ident->intrinsic() == semantic::Intrinsic::kPack2x16Snorm) { + if (ident->intrinsic() == semantic::IntrinsicType::kPack4x8Snorm || + ident->intrinsic() == semantic::IntrinsicType::kPack2x16Snorm) { is_signed = true; scale = (scale - 1) / 2; } switch (ident->intrinsic()) { - case semantic::Intrinsic::kPack4x8Snorm: - case semantic::Intrinsic::kPack4x8Unorm: - case semantic::Intrinsic::kPack2x16Snorm: - case semantic::Intrinsic::kPack2x16Unorm: + case semantic::IntrinsicType::kPack4x8Snorm: + case semantic::IntrinsicType::kPack4x8Unorm: + case semantic::IntrinsicType::kPack2x16Snorm: + case semantic::IntrinsicType::kPack2x16Unorm: pre << (is_signed ? "" : "u") << "int" << dims << " " << tmp_name << " = " << (is_signed ? "" : "u") << "int" << dims << "(round(clamp(" << expr_out.str() << ", " << (is_signed ? "-1.0" : "0.0") @@ -683,7 +683,7 @@ bool GeneratorImpl::EmitDataPackingCall(std::ostream& pre, } out << ")"; break; - case semantic::Intrinsic::kPack2x16Float: + case semantic::IntrinsicType::kPack2x16Float: pre << "uint2 " << tmp_name << " = f32tof16(" << expr_out.str() << ");\n"; out << "(" << tmp_name << ".x | " << tmp_name << ".y << 16)"; break; @@ -709,17 +709,17 @@ bool GeneratorImpl::EmitTextureCall(std::ostream& pre, auto* texture_type = TypeOf(texture)->UnwrapAll()->As(); switch (sem->intrinsic()) { - case semantic::Intrinsic::kTextureDimensions: - case semantic::Intrinsic::kTextureNumLayers: - case semantic::Intrinsic::kTextureNumLevels: - case semantic::Intrinsic::kTextureNumSamples: { + case semantic::IntrinsicType::kTextureDimensions: + case semantic::IntrinsicType::kTextureNumLayers: + case semantic::IntrinsicType::kTextureNumLevels: + case semantic::IntrinsicType::kTextureNumSamples: { // All of these intrinsics use the GetDimensions() method on the texture int num_dimensions = 0; const char* swizzle = ""; bool add_mip_level_in = false; switch (sem->intrinsic()) { - case semantic::Intrinsic::kTextureDimensions: + case semantic::IntrinsicType::kTextureDimensions: switch (texture_type->dim()) { case type::TextureDimension::kNone: error_ = "texture dimension is kNone"; @@ -755,7 +755,7 @@ bool GeneratorImpl::EmitTextureCall(std::ostream& pre, break; } break; - case semantic::Intrinsic::kTextureNumLayers: + case semantic::IntrinsicType::kTextureNumLayers: switch (texture_type->dim()) { default: error_ = "texture dimension is not arrayed"; @@ -771,7 +771,7 @@ bool GeneratorImpl::EmitTextureCall(std::ostream& pre, break; } break; - case semantic::Intrinsic::kTextureNumLevels: + case semantic::IntrinsicType::kTextureNumLevels: add_mip_level_in = true; switch (texture_type->dim()) { default: @@ -790,7 +790,7 @@ bool GeneratorImpl::EmitTextureCall(std::ostream& pre, break; } break; - case semantic::Intrinsic::kTextureNumSamples: + case semantic::IntrinsicType::kTextureNumSamples: switch (texture_type->dim()) { default: error_ = "texture dimension does not support multisampling"; @@ -860,28 +860,28 @@ bool GeneratorImpl::EmitTextureCall(std::ostream& pre, bool pack_mip_in_coords = false; switch (sem->intrinsic()) { - case semantic::Intrinsic::kTextureSample: + case semantic::IntrinsicType::kTextureSample: out << ".Sample("; break; - case semantic::Intrinsic::kTextureSampleBias: + case semantic::IntrinsicType::kTextureSampleBias: out << ".SampleBias("; break; - case semantic::Intrinsic::kTextureSampleLevel: + case semantic::IntrinsicType::kTextureSampleLevel: out << ".SampleLevel("; break; - case semantic::Intrinsic::kTextureSampleGrad: + case semantic::IntrinsicType::kTextureSampleGrad: out << ".SampleGrad("; break; - case semantic::Intrinsic::kTextureSampleCompare: + case semantic::IntrinsicType::kTextureSampleCompare: out << ".SampleCmp("; break; - case semantic::Intrinsic::kTextureLoad: + case semantic::IntrinsicType::kTextureLoad: out << ".Load("; if (!texture_type->Is()) { pack_mip_in_coords = true; } break; - case semantic::Intrinsic::kTextureStore: + case semantic::IntrinsicType::kTextureStore: out << "["; break; default: @@ -937,7 +937,7 @@ bool GeneratorImpl::EmitTextureCall(std::ostream& pre, } } - if (sem->intrinsic() == semantic::Intrinsic::kTextureStore) { + if (sem->intrinsic() == semantic::IntrinsicType::kTextureStore) { out << "] = "; if (!EmitExpression(pre, out, params[pidx.value])) return false; @@ -952,94 +952,94 @@ std::string GeneratorImpl::generate_builtin_name( const semantic::IntrinsicCall* call) { std::string out; switch (call->intrinsic()) { - case semantic::Intrinsic::kAcos: - case semantic::Intrinsic::kAny: - case semantic::Intrinsic::kAll: - case semantic::Intrinsic::kAsin: - case semantic::Intrinsic::kAtan: - case semantic::Intrinsic::kAtan2: - case semantic::Intrinsic::kCeil: - case semantic::Intrinsic::kCos: - case semantic::Intrinsic::kCosh: - case semantic::Intrinsic::kCross: - case semantic::Intrinsic::kDeterminant: - case semantic::Intrinsic::kDistance: - case semantic::Intrinsic::kDot: - case semantic::Intrinsic::kExp: - case semantic::Intrinsic::kExp2: - case semantic::Intrinsic::kFloor: - case semantic::Intrinsic::kFma: - case semantic::Intrinsic::kLdexp: - case semantic::Intrinsic::kLength: - case semantic::Intrinsic::kLog: - case semantic::Intrinsic::kLog2: - case semantic::Intrinsic::kNormalize: - case semantic::Intrinsic::kPow: - case semantic::Intrinsic::kReflect: - case semantic::Intrinsic::kRound: - case semantic::Intrinsic::kSin: - case semantic::Intrinsic::kSinh: - case semantic::Intrinsic::kSqrt: - case semantic::Intrinsic::kStep: - case semantic::Intrinsic::kTan: - case semantic::Intrinsic::kTanh: - case semantic::Intrinsic::kTrunc: - case semantic::Intrinsic::kMix: - case semantic::Intrinsic::kSign: - case semantic::Intrinsic::kAbs: - case semantic::Intrinsic::kMax: - case semantic::Intrinsic::kMin: - case semantic::Intrinsic::kClamp: + case semantic::IntrinsicType::kAcos: + case semantic::IntrinsicType::kAny: + case semantic::IntrinsicType::kAll: + case semantic::IntrinsicType::kAsin: + case semantic::IntrinsicType::kAtan: + case semantic::IntrinsicType::kAtan2: + case semantic::IntrinsicType::kCeil: + case semantic::IntrinsicType::kCos: + case semantic::IntrinsicType::kCosh: + case semantic::IntrinsicType::kCross: + case semantic::IntrinsicType::kDeterminant: + case semantic::IntrinsicType::kDistance: + case semantic::IntrinsicType::kDot: + case semantic::IntrinsicType::kExp: + case semantic::IntrinsicType::kExp2: + case semantic::IntrinsicType::kFloor: + case semantic::IntrinsicType::kFma: + case semantic::IntrinsicType::kLdexp: + case semantic::IntrinsicType::kLength: + case semantic::IntrinsicType::kLog: + case semantic::IntrinsicType::kLog2: + case semantic::IntrinsicType::kNormalize: + case semantic::IntrinsicType::kPow: + case semantic::IntrinsicType::kReflect: + case semantic::IntrinsicType::kRound: + case semantic::IntrinsicType::kSin: + case semantic::IntrinsicType::kSinh: + case semantic::IntrinsicType::kSqrt: + case semantic::IntrinsicType::kStep: + case semantic::IntrinsicType::kTan: + case semantic::IntrinsicType::kTanh: + case semantic::IntrinsicType::kTrunc: + case semantic::IntrinsicType::kMix: + case semantic::IntrinsicType::kSign: + case semantic::IntrinsicType::kAbs: + case semantic::IntrinsicType::kMax: + case semantic::IntrinsicType::kMin: + case semantic::IntrinsicType::kClamp: out = semantic::intrinsic::str(call->intrinsic()); break; - case semantic::Intrinsic::kCountOneBits: + case semantic::IntrinsicType::kCountOneBits: out = "countbits"; break; - case semantic::Intrinsic::kDpdx: + case semantic::IntrinsicType::kDpdx: out = "ddx"; break; - case semantic::Intrinsic::kDpdxCoarse: + case semantic::IntrinsicType::kDpdxCoarse: out = "ddx_coarse"; break; - case semantic::Intrinsic::kDpdxFine: + case semantic::IntrinsicType::kDpdxFine: out = "ddx_fine"; break; - case semantic::Intrinsic::kDpdy: + case semantic::IntrinsicType::kDpdy: out = "ddy"; break; - case semantic::Intrinsic::kDpdyCoarse: + case semantic::IntrinsicType::kDpdyCoarse: out = "ddy_coarse"; break; - case semantic::Intrinsic::kDpdyFine: + case semantic::IntrinsicType::kDpdyFine: out = "ddy_fine"; break; - case semantic::Intrinsic::kFaceForward: + case semantic::IntrinsicType::kFaceForward: out = "faceforward"; break; - case semantic::Intrinsic::kFract: + case semantic::IntrinsicType::kFract: out = "frac"; break; - case semantic::Intrinsic::kFwidth: - case semantic::Intrinsic::kFwidthCoarse: - case semantic::Intrinsic::kFwidthFine: + case semantic::IntrinsicType::kFwidth: + case semantic::IntrinsicType::kFwidthCoarse: + case semantic::IntrinsicType::kFwidthFine: out = "fwidth"; break; - case semantic::Intrinsic::kInverseSqrt: + case semantic::IntrinsicType::kInverseSqrt: out = "rsqrt"; break; - case semantic::Intrinsic::kIsFinite: + case semantic::IntrinsicType::kIsFinite: out = "isfinite"; break; - case semantic::Intrinsic::kIsInf: + case semantic::IntrinsicType::kIsInf: out = "isinf"; break; - case semantic::Intrinsic::kIsNan: + case semantic::IntrinsicType::kIsNan: out = "isnan"; break; - case semantic::Intrinsic::kReverseBits: + case semantic::IntrinsicType::kReverseBits: out = "reversebits"; break; - case semantic::Intrinsic::kSmoothStep: + case semantic::IntrinsicType::kSmoothStep: out = "smoothstep"; break; default: diff --git a/src/writer/hlsl/generator_impl_intrinsic_test.cc b/src/writer/hlsl/generator_impl_intrinsic_test.cc index 1976462dae..27a76e5d41 100644 --- a/src/writer/hlsl/generator_impl_intrinsic_test.cc +++ b/src/writer/hlsl/generator_impl_intrinsic_test.cc @@ -29,6 +29,8 @@ namespace writer { namespace hlsl { namespace { +using IntrinsicType = semantic::IntrinsicType; + using ::testing::HasSubstr; using HlslGeneratorImplTest_Intrinsic = TestHelper; @@ -40,7 +42,7 @@ enum class ParamType { }; struct IntrinsicData { - semantic::Intrinsic intrinsic; + IntrinsicType intrinsic; ParamType type; const char* hlsl_name; }; @@ -61,92 +63,92 @@ inline std::ostream& operator<<(std::ostream& out, IntrinsicData data) { return out; } -ast::CallExpression* GenerateCall(semantic::Intrinsic intrinsic, +ast::CallExpression* GenerateCall(IntrinsicType intrinsic, ParamType type, ProgramBuilder* builder) { std::string name; std::ostringstream str(name); str << intrinsic; switch (intrinsic) { - case semantic::Intrinsic::kAcos: - case semantic::Intrinsic::kAsin: - case semantic::Intrinsic::kAtan: - case semantic::Intrinsic::kCeil: - case semantic::Intrinsic::kCos: - case semantic::Intrinsic::kCosh: - case semantic::Intrinsic::kDpdx: - case semantic::Intrinsic::kDpdxCoarse: - case semantic::Intrinsic::kDpdxFine: - case semantic::Intrinsic::kDpdy: - case semantic::Intrinsic::kDpdyCoarse: - case semantic::Intrinsic::kDpdyFine: - case semantic::Intrinsic::kExp: - case semantic::Intrinsic::kExp2: - case semantic::Intrinsic::kFloor: - case semantic::Intrinsic::kFract: - case semantic::Intrinsic::kFwidth: - case semantic::Intrinsic::kFwidthCoarse: - case semantic::Intrinsic::kFwidthFine: - case semantic::Intrinsic::kInverseSqrt: - case semantic::Intrinsic::kIsFinite: - case semantic::Intrinsic::kIsInf: - case semantic::Intrinsic::kIsNan: - case semantic::Intrinsic::kIsNormal: - case semantic::Intrinsic::kLdexp: - case semantic::Intrinsic::kLength: - case semantic::Intrinsic::kLog: - case semantic::Intrinsic::kLog2: - case semantic::Intrinsic::kNormalize: - case semantic::Intrinsic::kReflect: - case semantic::Intrinsic::kRound: - case semantic::Intrinsic::kSin: - case semantic::Intrinsic::kSinh: - case semantic::Intrinsic::kSqrt: - case semantic::Intrinsic::kTan: - case semantic::Intrinsic::kTanh: - case semantic::Intrinsic::kTrunc: - case semantic::Intrinsic::kSign: + case IntrinsicType::kAcos: + case IntrinsicType::kAsin: + case IntrinsicType::kAtan: + case IntrinsicType::kCeil: + case IntrinsicType::kCos: + case IntrinsicType::kCosh: + case IntrinsicType::kDpdx: + case IntrinsicType::kDpdxCoarse: + case IntrinsicType::kDpdxFine: + case IntrinsicType::kDpdy: + case IntrinsicType::kDpdyCoarse: + case IntrinsicType::kDpdyFine: + case IntrinsicType::kExp: + case IntrinsicType::kExp2: + case IntrinsicType::kFloor: + case IntrinsicType::kFract: + case IntrinsicType::kFwidth: + case IntrinsicType::kFwidthCoarse: + case IntrinsicType::kFwidthFine: + case IntrinsicType::kInverseSqrt: + case IntrinsicType::kIsFinite: + case IntrinsicType::kIsInf: + case IntrinsicType::kIsNan: + case IntrinsicType::kIsNormal: + case IntrinsicType::kLdexp: + case IntrinsicType::kLength: + case IntrinsicType::kLog: + case IntrinsicType::kLog2: + case IntrinsicType::kNormalize: + case IntrinsicType::kReflect: + case IntrinsicType::kRound: + case IntrinsicType::kSin: + case IntrinsicType::kSinh: + case IntrinsicType::kSqrt: + case IntrinsicType::kTan: + case IntrinsicType::kTanh: + case IntrinsicType::kTrunc: + case IntrinsicType::kSign: return builder->Call(str.str(), "f1"); - case semantic::Intrinsic::kAtan2: - case semantic::Intrinsic::kCross: - case semantic::Intrinsic::kDot: - case semantic::Intrinsic::kDistance: - case semantic::Intrinsic::kPow: - case semantic::Intrinsic::kStep: + case IntrinsicType::kAtan2: + case IntrinsicType::kCross: + case IntrinsicType::kDot: + case IntrinsicType::kDistance: + case IntrinsicType::kPow: + case IntrinsicType::kStep: return builder->Call(str.str(), "f1", "f2"); - case semantic::Intrinsic::kFma: - case semantic::Intrinsic::kMix: - case semantic::Intrinsic::kFaceForward: - case semantic::Intrinsic::kSmoothStep: + case IntrinsicType::kFma: + case IntrinsicType::kMix: + case IntrinsicType::kFaceForward: + case IntrinsicType::kSmoothStep: return builder->Call(str.str(), "f1", "f2", "f3"); - case semantic::Intrinsic::kAll: - case semantic::Intrinsic::kAny: + case IntrinsicType::kAll: + case IntrinsicType::kAny: return builder->Call(str.str(), "b1"); - case semantic::Intrinsic::kAbs: + case IntrinsicType::kAbs: if (type == ParamType::kF32) { return builder->Call(str.str(), "f1"); } else { return builder->Call(str.str(), "u1"); } - case semantic::Intrinsic::kCountOneBits: - case semantic::Intrinsic::kReverseBits: + case IntrinsicType::kCountOneBits: + case IntrinsicType::kReverseBits: return builder->Call(str.str(), "u1"); - case semantic::Intrinsic::kMax: - case semantic::Intrinsic::kMin: + case IntrinsicType::kMax: + case IntrinsicType::kMin: if (type == ParamType::kF32) { return builder->Call(str.str(), "f1", "f2"); } else { return builder->Call(str.str(), "u1", "u2"); } - case semantic::Intrinsic::kClamp: + case IntrinsicType::kClamp: if (type == ParamType::kF32) { return builder->Call(str.str(), "f1", "f2", "f3"); } else { return builder->Call(str.str(), "u1", "u2", "u3"); } - case semantic::Intrinsic::kSelect: + case IntrinsicType::kSelect: return builder->Call(str.str(), "f1", "f2", "b1"); - case semantic::Intrinsic::kDeterminant: + case IntrinsicType::kDeterminant: return builder->Call(str.str(), "m1"); default: break; @@ -183,81 +185,72 @@ INSTANTIATE_TEST_SUITE_P( HlslGeneratorImplTest_Intrinsic, HlslIntrinsicTest, testing::Values( - IntrinsicData{semantic::Intrinsic::kAbs, ParamType::kF32, "abs"}, - IntrinsicData{semantic::Intrinsic::kAbs, ParamType::kU32, "abs"}, - IntrinsicData{semantic::Intrinsic::kAcos, ParamType::kF32, "acos"}, - IntrinsicData{semantic::Intrinsic::kAll, ParamType::kBool, "all"}, - IntrinsicData{semantic::Intrinsic::kAny, ParamType::kBool, "any"}, - IntrinsicData{semantic::Intrinsic::kAsin, ParamType::kF32, "asin"}, - IntrinsicData{semantic::Intrinsic::kAtan, ParamType::kF32, "atan"}, - IntrinsicData{semantic::Intrinsic::kAtan2, ParamType::kF32, "atan2"}, - IntrinsicData{semantic::Intrinsic::kCeil, ParamType::kF32, "ceil"}, - IntrinsicData{semantic::Intrinsic::kClamp, ParamType::kF32, "clamp"}, - IntrinsicData{semantic::Intrinsic::kClamp, ParamType::kU32, "clamp"}, - IntrinsicData{semantic::Intrinsic::kCos, ParamType::kF32, "cos"}, - IntrinsicData{semantic::Intrinsic::kCosh, ParamType::kF32, "cosh"}, - IntrinsicData{semantic::Intrinsic::kCountOneBits, ParamType::kU32, + IntrinsicData{IntrinsicType::kAbs, ParamType::kF32, "abs"}, + IntrinsicData{IntrinsicType::kAbs, ParamType::kU32, "abs"}, + IntrinsicData{IntrinsicType::kAcos, ParamType::kF32, "acos"}, + IntrinsicData{IntrinsicType::kAll, ParamType::kBool, "all"}, + IntrinsicData{IntrinsicType::kAny, ParamType::kBool, "any"}, + IntrinsicData{IntrinsicType::kAsin, ParamType::kF32, "asin"}, + IntrinsicData{IntrinsicType::kAtan, ParamType::kF32, "atan"}, + IntrinsicData{IntrinsicType::kAtan2, ParamType::kF32, "atan2"}, + IntrinsicData{IntrinsicType::kCeil, ParamType::kF32, "ceil"}, + IntrinsicData{IntrinsicType::kClamp, ParamType::kF32, "clamp"}, + IntrinsicData{IntrinsicType::kClamp, ParamType::kU32, "clamp"}, + IntrinsicData{IntrinsicType::kCos, ParamType::kF32, "cos"}, + IntrinsicData{IntrinsicType::kCosh, ParamType::kF32, "cosh"}, + IntrinsicData{IntrinsicType::kCountOneBits, ParamType::kU32, "countbits"}, - IntrinsicData{semantic::Intrinsic::kCross, ParamType::kF32, "cross"}, - IntrinsicData{semantic::Intrinsic::kDeterminant, ParamType::kF32, + IntrinsicData{IntrinsicType::kCross, ParamType::kF32, "cross"}, + IntrinsicData{IntrinsicType::kDeterminant, ParamType::kF32, "determinant"}, - IntrinsicData{semantic::Intrinsic::kDistance, ParamType::kF32, - "distance"}, - IntrinsicData{semantic::Intrinsic::kDot, ParamType::kF32, "dot"}, - IntrinsicData{semantic::Intrinsic::kDpdx, ParamType::kF32, "ddx"}, - IntrinsicData{semantic::Intrinsic::kDpdxCoarse, ParamType::kF32, + IntrinsicData{IntrinsicType::kDistance, ParamType::kF32, "distance"}, + IntrinsicData{IntrinsicType::kDot, ParamType::kF32, "dot"}, + IntrinsicData{IntrinsicType::kDpdx, ParamType::kF32, "ddx"}, + IntrinsicData{IntrinsicType::kDpdxCoarse, ParamType::kF32, "ddx_coarse"}, - IntrinsicData{semantic::Intrinsic::kDpdxFine, ParamType::kF32, - "ddx_fine"}, - IntrinsicData{semantic::Intrinsic::kDpdy, ParamType::kF32, "ddy"}, - IntrinsicData{semantic::Intrinsic::kDpdyCoarse, ParamType::kF32, + IntrinsicData{IntrinsicType::kDpdxFine, ParamType::kF32, "ddx_fine"}, + IntrinsicData{IntrinsicType::kDpdy, ParamType::kF32, "ddy"}, + IntrinsicData{IntrinsicType::kDpdyCoarse, ParamType::kF32, "ddy_coarse"}, - IntrinsicData{semantic::Intrinsic::kDpdyFine, ParamType::kF32, - "ddy_fine"}, - IntrinsicData{semantic::Intrinsic::kExp, ParamType::kF32, "exp"}, - IntrinsicData{semantic::Intrinsic::kExp2, ParamType::kF32, "exp2"}, - IntrinsicData{semantic::Intrinsic::kFaceForward, ParamType::kF32, + IntrinsicData{IntrinsicType::kDpdyFine, ParamType::kF32, "ddy_fine"}, + IntrinsicData{IntrinsicType::kExp, ParamType::kF32, "exp"}, + IntrinsicData{IntrinsicType::kExp2, ParamType::kF32, "exp2"}, + IntrinsicData{IntrinsicType::kFaceForward, ParamType::kF32, "faceforward"}, - IntrinsicData{semantic::Intrinsic::kFloor, ParamType::kF32, "floor"}, - IntrinsicData{semantic::Intrinsic::kFma, ParamType::kF32, "fma"}, - IntrinsicData{semantic::Intrinsic::kFract, ParamType::kF32, "frac"}, - IntrinsicData{semantic::Intrinsic::kFwidth, ParamType::kF32, "fwidth"}, - IntrinsicData{semantic::Intrinsic::kFwidthCoarse, ParamType::kF32, - "fwidth"}, - IntrinsicData{semantic::Intrinsic::kFwidthFine, ParamType::kF32, - "fwidth"}, - IntrinsicData{semantic::Intrinsic::kInverseSqrt, ParamType::kF32, - "rsqrt"}, - IntrinsicData{semantic::Intrinsic::kIsFinite, ParamType::kF32, - "isfinite"}, - IntrinsicData{semantic::Intrinsic::kIsInf, ParamType::kF32, "isinf"}, - IntrinsicData{semantic::Intrinsic::kIsNan, ParamType::kF32, "isnan"}, - IntrinsicData{semantic::Intrinsic::kLdexp, ParamType::kF32, "ldexp"}, - IntrinsicData{semantic::Intrinsic::kLength, ParamType::kF32, "length"}, - IntrinsicData{semantic::Intrinsic::kLog, ParamType::kF32, "log"}, - IntrinsicData{semantic::Intrinsic::kLog2, ParamType::kF32, "log2"}, - IntrinsicData{semantic::Intrinsic::kMax, ParamType::kF32, "max"}, - IntrinsicData{semantic::Intrinsic::kMax, ParamType::kU32, "max"}, - IntrinsicData{semantic::Intrinsic::kMin, ParamType::kF32, "min"}, - IntrinsicData{semantic::Intrinsic::kMin, ParamType::kU32, "min"}, - IntrinsicData{semantic::Intrinsic::kNormalize, ParamType::kF32, - "normalize"}, - IntrinsicData{semantic::Intrinsic::kPow, ParamType::kF32, "pow"}, - IntrinsicData{semantic::Intrinsic::kReflect, ParamType::kF32, - "reflect"}, - IntrinsicData{semantic::Intrinsic::kReverseBits, ParamType::kU32, + IntrinsicData{IntrinsicType::kFloor, ParamType::kF32, "floor"}, + IntrinsicData{IntrinsicType::kFma, ParamType::kF32, "fma"}, + IntrinsicData{IntrinsicType::kFract, ParamType::kF32, "frac"}, + IntrinsicData{IntrinsicType::kFwidth, ParamType::kF32, "fwidth"}, + IntrinsicData{IntrinsicType::kFwidthCoarse, ParamType::kF32, "fwidth"}, + IntrinsicData{IntrinsicType::kFwidthFine, ParamType::kF32, "fwidth"}, + IntrinsicData{IntrinsicType::kInverseSqrt, ParamType::kF32, "rsqrt"}, + IntrinsicData{IntrinsicType::kIsFinite, ParamType::kF32, "isfinite"}, + IntrinsicData{IntrinsicType::kIsInf, ParamType::kF32, "isinf"}, + IntrinsicData{IntrinsicType::kIsNan, ParamType::kF32, "isnan"}, + IntrinsicData{IntrinsicType::kLdexp, ParamType::kF32, "ldexp"}, + IntrinsicData{IntrinsicType::kLength, ParamType::kF32, "length"}, + IntrinsicData{IntrinsicType::kLog, ParamType::kF32, "log"}, + IntrinsicData{IntrinsicType::kLog2, ParamType::kF32, "log2"}, + IntrinsicData{IntrinsicType::kMax, ParamType::kF32, "max"}, + IntrinsicData{IntrinsicType::kMax, ParamType::kU32, "max"}, + IntrinsicData{IntrinsicType::kMin, ParamType::kF32, "min"}, + IntrinsicData{IntrinsicType::kMin, ParamType::kU32, "min"}, + IntrinsicData{IntrinsicType::kNormalize, ParamType::kF32, "normalize"}, + IntrinsicData{IntrinsicType::kPow, ParamType::kF32, "pow"}, + IntrinsicData{IntrinsicType::kReflect, ParamType::kF32, "reflect"}, + IntrinsicData{IntrinsicType::kReverseBits, ParamType::kU32, "reversebits"}, - IntrinsicData{semantic::Intrinsic::kRound, ParamType::kU32, "round"}, - IntrinsicData{semantic::Intrinsic::kSign, ParamType::kF32, "sign"}, - IntrinsicData{semantic::Intrinsic::kSin, ParamType::kF32, "sin"}, - IntrinsicData{semantic::Intrinsic::kSinh, ParamType::kF32, "sinh"}, - IntrinsicData{semantic::Intrinsic::kSmoothStep, ParamType::kF32, + IntrinsicData{IntrinsicType::kRound, ParamType::kU32, "round"}, + IntrinsicData{IntrinsicType::kSign, ParamType::kF32, "sign"}, + IntrinsicData{IntrinsicType::kSin, ParamType::kF32, "sin"}, + IntrinsicData{IntrinsicType::kSinh, ParamType::kF32, "sinh"}, + IntrinsicData{IntrinsicType::kSmoothStep, ParamType::kF32, "smoothstep"}, - IntrinsicData{semantic::Intrinsic::kSqrt, ParamType::kF32, "sqrt"}, - IntrinsicData{semantic::Intrinsic::kStep, ParamType::kF32, "step"}, - IntrinsicData{semantic::Intrinsic::kTan, ParamType::kF32, "tan"}, - IntrinsicData{semantic::Intrinsic::kTanh, ParamType::kF32, "tanh"}, - IntrinsicData{semantic::Intrinsic::kTrunc, ParamType::kF32, "trunc"})); + IntrinsicData{IntrinsicType::kSqrt, ParamType::kF32, "sqrt"}, + IntrinsicData{IntrinsicType::kStep, ParamType::kF32, "step"}, + IntrinsicData{IntrinsicType::kTan, ParamType::kF32, "tan"}, + IntrinsicData{IntrinsicType::kTanh, ParamType::kF32, "tanh"}, + IntrinsicData{IntrinsicType::kTrunc, ParamType::kF32, "trunc"})); TEST_F(HlslGeneratorImplTest_Intrinsic, DISABLED_Intrinsic_IsNormal) { FAIL(); diff --git a/src/writer/msl/generator_impl.cc b/src/writer/msl/generator_impl.cc index cdce7070c4..002050122e 100644 --- a/src/writer/msl/generator_impl.cc +++ b/src/writer/msl/generator_impl.cc @@ -449,7 +449,7 @@ bool GeneratorImpl::EmitCall(ast::CallExpression* expr) { return EmitTextureCall(expr, sem); } if (auto* sem = call_sem->As()) { - if (sem->intrinsic() == semantic::Intrinsic::kPack2x16Float) { + if (sem->intrinsic() == semantic::IntrinsicType::kPack2x16Float) { make_indent(); out_ << "as_type(half2("; if (!EmitExpression(expr->params()[0])) { @@ -580,7 +580,7 @@ bool GeneratorImpl::EmitTextureCall(ast::CallExpression* expr, TypeOf(params[pidx.texture])->UnwrapAll()->As(); switch (sem->intrinsic()) { - case semantic::Intrinsic::kTextureDimensions: { + case semantic::IntrinsicType::kTextureDimensions: { std::vector dims; switch (texture_type->dim()) { case type::TextureDimension::kNone: @@ -634,7 +634,7 @@ bool GeneratorImpl::EmitTextureCall(ast::CallExpression* expr, } return true; } - case semantic::Intrinsic::kTextureNumLayers: { + case semantic::IntrinsicType::kTextureNumLayers: { out_ << "int("; if (!EmitExpression(params[pidx.texture])) { return false; @@ -642,7 +642,7 @@ bool GeneratorImpl::EmitTextureCall(ast::CallExpression* expr, out_ << ".get_array_size())"; return true; } - case semantic::Intrinsic::kTextureNumLevels: { + case semantic::IntrinsicType::kTextureNumLevels: { out_ << "int("; if (!EmitExpression(params[pidx.texture])) { return false; @@ -650,7 +650,7 @@ bool GeneratorImpl::EmitTextureCall(ast::CallExpression* expr, out_ << ".get_num_mip_levels())"; return true; } - case semantic::Intrinsic::kTextureNumSamples: { + case semantic::IntrinsicType::kTextureNumSamples: { out_ << "int("; if (!EmitExpression(params[pidx.texture])) { return false; @@ -668,20 +668,20 @@ bool GeneratorImpl::EmitTextureCall(ast::CallExpression* expr, bool lod_param_is_named = true; switch (sem->intrinsic()) { - case semantic::Intrinsic::kTextureSample: - case semantic::Intrinsic::kTextureSampleBias: - case semantic::Intrinsic::kTextureSampleLevel: - case semantic::Intrinsic::kTextureSampleGrad: + case semantic::IntrinsicType::kTextureSample: + case semantic::IntrinsicType::kTextureSampleBias: + case semantic::IntrinsicType::kTextureSampleLevel: + case semantic::IntrinsicType::kTextureSampleGrad: out_ << ".sample("; break; - case semantic::Intrinsic::kTextureSampleCompare: + case semantic::IntrinsicType::kTextureSampleCompare: out_ << ".sample_compare("; break; - case semantic::Intrinsic::kTextureLoad: + case semantic::IntrinsicType::kTextureLoad: out_ << ".read("; lod_param_is_named = false; break; - case semantic::Intrinsic::kTextureStore: + case semantic::IntrinsicType::kTextureStore: out_ << ".write("; break; default: @@ -780,118 +780,118 @@ std::string GeneratorImpl::generate_builtin_name( const semantic::IntrinsicCall* call) { std::string out = "metal::"; switch (call->intrinsic()) { - case semantic::Intrinsic::kAcos: - case semantic::Intrinsic::kAll: - case semantic::Intrinsic::kAny: - case semantic::Intrinsic::kAsin: - case semantic::Intrinsic::kAtan: - case semantic::Intrinsic::kAtan2: - case semantic::Intrinsic::kCeil: - case semantic::Intrinsic::kCos: - case semantic::Intrinsic::kCosh: - case semantic::Intrinsic::kCross: - case semantic::Intrinsic::kDeterminant: - case semantic::Intrinsic::kDistance: - case semantic::Intrinsic::kDot: - case semantic::Intrinsic::kExp: - case semantic::Intrinsic::kExp2: - case semantic::Intrinsic::kFloor: - case semantic::Intrinsic::kFma: - case semantic::Intrinsic::kFract: - case semantic::Intrinsic::kLength: - case semantic::Intrinsic::kLdexp: - case semantic::Intrinsic::kLog: - case semantic::Intrinsic::kLog2: - case semantic::Intrinsic::kMix: - case semantic::Intrinsic::kNormalize: - case semantic::Intrinsic::kPow: - case semantic::Intrinsic::kReflect: - case semantic::Intrinsic::kRound: - case semantic::Intrinsic::kSelect: - case semantic::Intrinsic::kSin: - case semantic::Intrinsic::kSinh: - case semantic::Intrinsic::kSqrt: - case semantic::Intrinsic::kStep: - case semantic::Intrinsic::kTan: - case semantic::Intrinsic::kTanh: - case semantic::Intrinsic::kTrunc: - case semantic::Intrinsic::kSign: - case semantic::Intrinsic::kClamp: + case semantic::IntrinsicType::kAcos: + case semantic::IntrinsicType::kAll: + case semantic::IntrinsicType::kAny: + case semantic::IntrinsicType::kAsin: + case semantic::IntrinsicType::kAtan: + case semantic::IntrinsicType::kAtan2: + case semantic::IntrinsicType::kCeil: + case semantic::IntrinsicType::kCos: + case semantic::IntrinsicType::kCosh: + case semantic::IntrinsicType::kCross: + case semantic::IntrinsicType::kDeterminant: + case semantic::IntrinsicType::kDistance: + case semantic::IntrinsicType::kDot: + case semantic::IntrinsicType::kExp: + case semantic::IntrinsicType::kExp2: + case semantic::IntrinsicType::kFloor: + case semantic::IntrinsicType::kFma: + case semantic::IntrinsicType::kFract: + case semantic::IntrinsicType::kLength: + case semantic::IntrinsicType::kLdexp: + case semantic::IntrinsicType::kLog: + case semantic::IntrinsicType::kLog2: + case semantic::IntrinsicType::kMix: + case semantic::IntrinsicType::kNormalize: + case semantic::IntrinsicType::kPow: + case semantic::IntrinsicType::kReflect: + case semantic::IntrinsicType::kRound: + case semantic::IntrinsicType::kSelect: + case semantic::IntrinsicType::kSin: + case semantic::IntrinsicType::kSinh: + case semantic::IntrinsicType::kSqrt: + case semantic::IntrinsicType::kStep: + case semantic::IntrinsicType::kTan: + case semantic::IntrinsicType::kTanh: + case semantic::IntrinsicType::kTrunc: + case semantic::IntrinsicType::kSign: + case semantic::IntrinsicType::kClamp: out += semantic::intrinsic::str(call->intrinsic()); break; - case semantic::Intrinsic::kAbs: + case semantic::IntrinsicType::kAbs: if (call->Type()->is_float_scalar_or_vector()) { out += "fabs"; } else { out += "abs"; } break; - case semantic::Intrinsic::kCountOneBits: + case semantic::IntrinsicType::kCountOneBits: out += "popcount"; break; - case semantic::Intrinsic::kDpdx: - case semantic::Intrinsic::kDpdxCoarse: - case semantic::Intrinsic::kDpdxFine: + case semantic::IntrinsicType::kDpdx: + case semantic::IntrinsicType::kDpdxCoarse: + case semantic::IntrinsicType::kDpdxFine: out += "dfdx"; break; - case semantic::Intrinsic::kDpdy: - case semantic::Intrinsic::kDpdyCoarse: - case semantic::Intrinsic::kDpdyFine: + case semantic::IntrinsicType::kDpdy: + case semantic::IntrinsicType::kDpdyCoarse: + case semantic::IntrinsicType::kDpdyFine: out += "dfdy"; break; - case semantic::Intrinsic::kFwidth: - case semantic::Intrinsic::kFwidthCoarse: - case semantic::Intrinsic::kFwidthFine: + case semantic::IntrinsicType::kFwidth: + case semantic::IntrinsicType::kFwidthCoarse: + case semantic::IntrinsicType::kFwidthFine: out += "fwidth"; break; - case semantic::Intrinsic::kIsFinite: + case semantic::IntrinsicType::kIsFinite: out += "isfinite"; break; - case semantic::Intrinsic::kIsInf: + case semantic::IntrinsicType::kIsInf: out += "isinf"; break; - case semantic::Intrinsic::kIsNan: + case semantic::IntrinsicType::kIsNan: out += "isnan"; break; - case semantic::Intrinsic::kIsNormal: + case semantic::IntrinsicType::kIsNormal: out += "isnormal"; break; - case semantic::Intrinsic::kMax: + case semantic::IntrinsicType::kMax: if (call->Type()->is_float_scalar_or_vector()) { out += "fmax"; } else { out += "max"; } break; - case semantic::Intrinsic::kMin: + case semantic::IntrinsicType::kMin: if (call->Type()->is_float_scalar_or_vector()) { out += "fmin"; } else { out += "min"; } break; - case semantic::Intrinsic::kFaceForward: + case semantic::IntrinsicType::kFaceForward: out += "faceforward"; break; - case semantic::Intrinsic::kPack4x8Snorm: + case semantic::IntrinsicType::kPack4x8Snorm: out += "pack_float_to_snorm4x8"; break; - case semantic::Intrinsic::kPack4x8Unorm: + case semantic::IntrinsicType::kPack4x8Unorm: out += "pack_float_to_unorm4x8"; break; - case semantic::Intrinsic::kPack2x16Snorm: + case semantic::IntrinsicType::kPack2x16Snorm: out += "pack_float_to_snorm2x16"; break; - case semantic::Intrinsic::kPack2x16Unorm: + case semantic::IntrinsicType::kPack2x16Unorm: out += "pack_float_to_unorm2x16"; break; - case semantic::Intrinsic::kReverseBits: + case semantic::IntrinsicType::kReverseBits: out += "reverse_bits"; break; - case semantic::Intrinsic::kSmoothStep: + case semantic::IntrinsicType::kSmoothStep: out += "smoothstep"; break; - case semantic::Intrinsic::kInverseSqrt: + case semantic::IntrinsicType::kInverseSqrt: out += "rsqrt"; break; default: diff --git a/src/writer/msl/generator_impl_intrinsic_test.cc b/src/writer/msl/generator_impl_intrinsic_test.cc index 75400438c4..25559044e1 100644 --- a/src/writer/msl/generator_impl_intrinsic_test.cc +++ b/src/writer/msl/generator_impl_intrinsic_test.cc @@ -30,6 +30,8 @@ namespace writer { namespace msl { namespace { +using IntrinsicType = semantic::IntrinsicType; + using MslGeneratorImplTest = TestHelper; enum class ParamType { @@ -39,7 +41,7 @@ enum class ParamType { }; struct IntrinsicData { - semantic::Intrinsic intrinsic; + IntrinsicType intrinsic; ParamType type; const char* msl_name; }; @@ -60,97 +62,97 @@ inline std::ostream& operator<<(std::ostream& out, IntrinsicData data) { return out; } -ast::CallExpression* GenerateCall(semantic::Intrinsic intrinsic, +ast::CallExpression* GenerateCall(IntrinsicType intrinsic, ParamType type, ProgramBuilder* builder) { std::string name; std::ostringstream str(name); str << intrinsic; switch (intrinsic) { - case semantic::Intrinsic::kAcos: - case semantic::Intrinsic::kAsin: - case semantic::Intrinsic::kAtan: - case semantic::Intrinsic::kCeil: - case semantic::Intrinsic::kCos: - case semantic::Intrinsic::kCosh: - case semantic::Intrinsic::kDpdx: - case semantic::Intrinsic::kDpdxCoarse: - case semantic::Intrinsic::kDpdxFine: - case semantic::Intrinsic::kDpdy: - case semantic::Intrinsic::kDpdyCoarse: - case semantic::Intrinsic::kDpdyFine: - case semantic::Intrinsic::kExp: - case semantic::Intrinsic::kExp2: - case semantic::Intrinsic::kFloor: - case semantic::Intrinsic::kFract: - case semantic::Intrinsic::kFwidth: - case semantic::Intrinsic::kFwidthCoarse: - case semantic::Intrinsic::kFwidthFine: - case semantic::Intrinsic::kInverseSqrt: - case semantic::Intrinsic::kIsFinite: - case semantic::Intrinsic::kIsInf: - case semantic::Intrinsic::kIsNan: - case semantic::Intrinsic::kIsNormal: - case semantic::Intrinsic::kLdexp: - case semantic::Intrinsic::kLength: - case semantic::Intrinsic::kLog: - case semantic::Intrinsic::kLog2: - case semantic::Intrinsic::kNormalize: - case semantic::Intrinsic::kPack4x8Snorm: - case semantic::Intrinsic::kPack4x8Unorm: - case semantic::Intrinsic::kReflect: - case semantic::Intrinsic::kRound: - case semantic::Intrinsic::kSin: - case semantic::Intrinsic::kSinh: - case semantic::Intrinsic::kSqrt: - case semantic::Intrinsic::kTan: - case semantic::Intrinsic::kTanh: - case semantic::Intrinsic::kTrunc: - case semantic::Intrinsic::kSign: + case IntrinsicType::kAcos: + case IntrinsicType::kAsin: + case IntrinsicType::kAtan: + case IntrinsicType::kCeil: + case IntrinsicType::kCos: + case IntrinsicType::kCosh: + case IntrinsicType::kDpdx: + case IntrinsicType::kDpdxCoarse: + case IntrinsicType::kDpdxFine: + case IntrinsicType::kDpdy: + case IntrinsicType::kDpdyCoarse: + case IntrinsicType::kDpdyFine: + case IntrinsicType::kExp: + case IntrinsicType::kExp2: + case IntrinsicType::kFloor: + case IntrinsicType::kFract: + case IntrinsicType::kFwidth: + case IntrinsicType::kFwidthCoarse: + case IntrinsicType::kFwidthFine: + case IntrinsicType::kInverseSqrt: + case IntrinsicType::kIsFinite: + case IntrinsicType::kIsInf: + case IntrinsicType::kIsNan: + case IntrinsicType::kIsNormal: + case IntrinsicType::kLdexp: + case IntrinsicType::kLength: + case IntrinsicType::kLog: + case IntrinsicType::kLog2: + case IntrinsicType::kNormalize: + case IntrinsicType::kPack4x8Snorm: + case IntrinsicType::kPack4x8Unorm: + case IntrinsicType::kReflect: + case IntrinsicType::kRound: + case IntrinsicType::kSin: + case IntrinsicType::kSinh: + case IntrinsicType::kSqrt: + case IntrinsicType::kTan: + case IntrinsicType::kTanh: + case IntrinsicType::kTrunc: + case IntrinsicType::kSign: return builder->Call(str.str(), "f1"); - case semantic::Intrinsic::kAtan2: - case semantic::Intrinsic::kCross: - case semantic::Intrinsic::kDot: - case semantic::Intrinsic::kDistance: - case semantic::Intrinsic::kPow: - case semantic::Intrinsic::kStep: + case IntrinsicType::kAtan2: + case IntrinsicType::kCross: + case IntrinsicType::kDot: + case IntrinsicType::kDistance: + case IntrinsicType::kPow: + case IntrinsicType::kStep: return builder->Call(str.str(), "f1", "f2"); - case semantic::Intrinsic::kFma: - case semantic::Intrinsic::kMix: - case semantic::Intrinsic::kFaceForward: - case semantic::Intrinsic::kSmoothStep: + case IntrinsicType::kFma: + case IntrinsicType::kMix: + case IntrinsicType::kFaceForward: + case IntrinsicType::kSmoothStep: return builder->Call(str.str(), "f1", "f2", "f3"); - case semantic::Intrinsic::kAll: - case semantic::Intrinsic::kAny: + case IntrinsicType::kAll: + case IntrinsicType::kAny: return builder->Call(str.str(), "b1"); - case semantic::Intrinsic::kAbs: + case IntrinsicType::kAbs: if (type == ParamType::kF32) { return builder->Call(str.str(), "f1"); } else { return builder->Call(str.str(), "u1"); } - case semantic::Intrinsic::kCountOneBits: - case semantic::Intrinsic::kReverseBits: + case IntrinsicType::kCountOneBits: + case IntrinsicType::kReverseBits: return builder->Call(str.str(), "u1"); - case semantic::Intrinsic::kMax: - case semantic::Intrinsic::kMin: + case IntrinsicType::kMax: + case IntrinsicType::kMin: if (type == ParamType::kF32) { return builder->Call(str.str(), "f1", "f2"); } else { return builder->Call(str.str(), "u1", "u2"); } - case semantic::Intrinsic::kClamp: + case IntrinsicType::kClamp: if (type == ParamType::kF32) { return builder->Call(str.str(), "f1", "f2", "f3"); } else { return builder->Call(str.str(), "u1", "u2", "u3"); } - case semantic::Intrinsic::kSelect: + case IntrinsicType::kSelect: return builder->Call(str.str(), "f1", "f2", "b1"); - case semantic::Intrinsic::kDeterminant: + case IntrinsicType::kDeterminant: return builder->Call(str.str(), "m1"); - case semantic::Intrinsic::kPack2x16Snorm: - case semantic::Intrinsic::kPack2x16Unorm: + case IntrinsicType::kPack2x16Snorm: + case IntrinsicType::kPack2x16Unorm: return builder->Call(str.str(), "f4"); default: break; @@ -189,125 +191,90 @@ INSTANTIATE_TEST_SUITE_P( MslGeneratorImplTest, MslIntrinsicTest, testing::Values( - IntrinsicData{semantic::Intrinsic::kAbs, ParamType::kF32, - "metal::fabs"}, - IntrinsicData{semantic::Intrinsic::kAbs, ParamType::kU32, "metal::abs"}, - IntrinsicData{semantic::Intrinsic::kAcos, ParamType::kF32, - "metal::acos"}, - IntrinsicData{semantic::Intrinsic::kAll, ParamType::kBool, - "metal::all"}, - IntrinsicData{semantic::Intrinsic::kAny, ParamType::kBool, - "metal::any"}, - IntrinsicData{semantic::Intrinsic::kAsin, ParamType::kF32, - "metal::asin"}, - IntrinsicData{semantic::Intrinsic::kAtan, ParamType::kF32, - "metal::atan"}, - IntrinsicData{semantic::Intrinsic::kAtan2, ParamType::kF32, - "metal::atan2"}, - IntrinsicData{semantic::Intrinsic::kCeil, ParamType::kF32, - "metal::ceil"}, - IntrinsicData{semantic::Intrinsic::kClamp, ParamType::kF32, - "metal::clamp"}, - IntrinsicData{semantic::Intrinsic::kClamp, ParamType::kU32, - "metal::clamp"}, - IntrinsicData{semantic::Intrinsic::kCos, ParamType::kF32, "metal::cos"}, - IntrinsicData{semantic::Intrinsic::kCosh, ParamType::kF32, - "metal::cosh"}, - IntrinsicData{semantic::Intrinsic::kCountOneBits, ParamType::kU32, + IntrinsicData{IntrinsicType::kAbs, ParamType::kF32, "metal::fabs"}, + IntrinsicData{IntrinsicType::kAbs, ParamType::kU32, "metal::abs"}, + IntrinsicData{IntrinsicType::kAcos, ParamType::kF32, "metal::acos"}, + IntrinsicData{IntrinsicType::kAll, ParamType::kBool, "metal::all"}, + IntrinsicData{IntrinsicType::kAny, ParamType::kBool, "metal::any"}, + IntrinsicData{IntrinsicType::kAsin, ParamType::kF32, "metal::asin"}, + IntrinsicData{IntrinsicType::kAtan, ParamType::kF32, "metal::atan"}, + IntrinsicData{IntrinsicType::kAtan2, ParamType::kF32, "metal::atan2"}, + IntrinsicData{IntrinsicType::kCeil, ParamType::kF32, "metal::ceil"}, + IntrinsicData{IntrinsicType::kClamp, ParamType::kF32, "metal::clamp"}, + IntrinsicData{IntrinsicType::kClamp, ParamType::kU32, "metal::clamp"}, + IntrinsicData{IntrinsicType::kCos, ParamType::kF32, "metal::cos"}, + IntrinsicData{IntrinsicType::kCosh, ParamType::kF32, "metal::cosh"}, + IntrinsicData{IntrinsicType::kCountOneBits, ParamType::kU32, "metal::popcount"}, - IntrinsicData{semantic::Intrinsic::kCross, ParamType::kF32, - "metal::cross"}, - IntrinsicData{semantic::Intrinsic::kDeterminant, ParamType::kF32, + IntrinsicData{IntrinsicType::kCross, ParamType::kF32, "metal::cross"}, + IntrinsicData{IntrinsicType::kDeterminant, ParamType::kF32, "metal::determinant"}, - IntrinsicData{semantic::Intrinsic::kDistance, ParamType::kF32, + IntrinsicData{IntrinsicType::kDistance, ParamType::kF32, "metal::distance"}, - IntrinsicData{semantic::Intrinsic::kDot, ParamType::kF32, "metal::dot"}, - IntrinsicData{semantic::Intrinsic::kDpdx, ParamType::kF32, + IntrinsicData{IntrinsicType::kDot, ParamType::kF32, "metal::dot"}, + IntrinsicData{IntrinsicType::kDpdx, ParamType::kF32, "metal::dfdx"}, + IntrinsicData{IntrinsicType::kDpdxCoarse, ParamType::kF32, "metal::dfdx"}, - IntrinsicData{semantic::Intrinsic::kDpdxCoarse, ParamType::kF32, - "metal::dfdx"}, - IntrinsicData{semantic::Intrinsic::kDpdxFine, ParamType::kF32, - "metal::dfdx"}, - IntrinsicData{semantic::Intrinsic::kDpdy, ParamType::kF32, + IntrinsicData{IntrinsicType::kDpdxFine, ParamType::kF32, "metal::dfdx"}, + IntrinsicData{IntrinsicType::kDpdy, ParamType::kF32, "metal::dfdy"}, + IntrinsicData{IntrinsicType::kDpdyCoarse, ParamType::kF32, "metal::dfdy"}, - IntrinsicData{semantic::Intrinsic::kDpdyCoarse, ParamType::kF32, - "metal::dfdy"}, - IntrinsicData{semantic::Intrinsic::kDpdyFine, ParamType::kF32, - "metal::dfdy"}, - IntrinsicData{semantic::Intrinsic::kExp, ParamType::kF32, "metal::exp"}, - IntrinsicData{semantic::Intrinsic::kExp2, ParamType::kF32, - "metal::exp2"}, - IntrinsicData{semantic::Intrinsic::kFaceForward, ParamType::kF32, + IntrinsicData{IntrinsicType::kDpdyFine, ParamType::kF32, "metal::dfdy"}, + IntrinsicData{IntrinsicType::kExp, ParamType::kF32, "metal::exp"}, + IntrinsicData{IntrinsicType::kExp2, ParamType::kF32, "metal::exp2"}, + IntrinsicData{IntrinsicType::kFaceForward, ParamType::kF32, "metal::faceforward"}, - IntrinsicData{semantic::Intrinsic::kFloor, ParamType::kF32, - "metal::floor"}, - IntrinsicData{semantic::Intrinsic::kFma, ParamType::kF32, "metal::fma"}, - IntrinsicData{semantic::Intrinsic::kFract, ParamType::kF32, - "metal::fract"}, - IntrinsicData{semantic::Intrinsic::kFwidth, ParamType::kF32, + IntrinsicData{IntrinsicType::kFloor, ParamType::kF32, "metal::floor"}, + IntrinsicData{IntrinsicType::kFma, ParamType::kF32, "metal::fma"}, + IntrinsicData{IntrinsicType::kFract, ParamType::kF32, "metal::fract"}, + IntrinsicData{IntrinsicType::kFwidth, ParamType::kF32, "metal::fwidth"}, + IntrinsicData{IntrinsicType::kFwidthCoarse, ParamType::kF32, "metal::fwidth"}, - IntrinsicData{semantic::Intrinsic::kFwidthCoarse, ParamType::kF32, + IntrinsicData{IntrinsicType::kFwidthFine, ParamType::kF32, "metal::fwidth"}, - IntrinsicData{semantic::Intrinsic::kFwidthFine, ParamType::kF32, - "metal::fwidth"}, - IntrinsicData{semantic::Intrinsic::kInverseSqrt, ParamType::kF32, + IntrinsicData{IntrinsicType::kInverseSqrt, ParamType::kF32, "metal::rsqrt"}, - IntrinsicData{semantic::Intrinsic::kIsFinite, ParamType::kF32, + IntrinsicData{IntrinsicType::kIsFinite, ParamType::kF32, "metal::isfinite"}, - IntrinsicData{semantic::Intrinsic::kIsInf, ParamType::kF32, - "metal::isinf"}, - IntrinsicData{semantic::Intrinsic::kIsNan, ParamType::kF32, - "metal::isnan"}, - IntrinsicData{semantic::Intrinsic::kIsNormal, ParamType::kF32, + IntrinsicData{IntrinsicType::kIsInf, ParamType::kF32, "metal::isinf"}, + IntrinsicData{IntrinsicType::kIsNan, ParamType::kF32, "metal::isnan"}, + IntrinsicData{IntrinsicType::kIsNormal, ParamType::kF32, "metal::isnormal"}, - IntrinsicData{semantic::Intrinsic::kLdexp, ParamType::kF32, - "metal::ldexp"}, - IntrinsicData{semantic::Intrinsic::kLength, ParamType::kF32, - "metal::length"}, - IntrinsicData{semantic::Intrinsic::kLog, ParamType::kF32, "metal::log"}, - IntrinsicData{semantic::Intrinsic::kLog2, ParamType::kF32, - "metal::log2"}, - IntrinsicData{semantic::Intrinsic::kMax, ParamType::kF32, - "metal::fmax"}, - IntrinsicData{semantic::Intrinsic::kMax, ParamType::kU32, "metal::max"}, - IntrinsicData{semantic::Intrinsic::kMin, ParamType::kF32, - "metal::fmin"}, - IntrinsicData{semantic::Intrinsic::kMin, ParamType::kU32, "metal::min"}, - IntrinsicData{semantic::Intrinsic::kNormalize, ParamType::kF32, + IntrinsicData{IntrinsicType::kLdexp, ParamType::kF32, "metal::ldexp"}, + IntrinsicData{IntrinsicType::kLength, ParamType::kF32, "metal::length"}, + IntrinsicData{IntrinsicType::kLog, ParamType::kF32, "metal::log"}, + IntrinsicData{IntrinsicType::kLog2, ParamType::kF32, "metal::log2"}, + IntrinsicData{IntrinsicType::kMax, ParamType::kF32, "metal::fmax"}, + IntrinsicData{IntrinsicType::kMax, ParamType::kU32, "metal::max"}, + IntrinsicData{IntrinsicType::kMin, ParamType::kF32, "metal::fmin"}, + IntrinsicData{IntrinsicType::kMin, ParamType::kU32, "metal::min"}, + IntrinsicData{IntrinsicType::kNormalize, ParamType::kF32, "metal::normalize"}, - IntrinsicData{semantic::Intrinsic::kPack4x8Snorm, ParamType::kF32, + IntrinsicData{IntrinsicType::kPack4x8Snorm, ParamType::kF32, "metal::pack_float_to_snorm4x8"}, - IntrinsicData{semantic::Intrinsic::kPack4x8Unorm, ParamType::kF32, + IntrinsicData{IntrinsicType::kPack4x8Unorm, ParamType::kF32, "metal::pack_float_to_unorm4x8"}, - IntrinsicData{semantic::Intrinsic::kPack2x16Snorm, ParamType::kF32, + IntrinsicData{IntrinsicType::kPack2x16Snorm, ParamType::kF32, "metal::pack_float_to_snorm2x16"}, - IntrinsicData{semantic::Intrinsic::kPack2x16Unorm, ParamType::kF32, + IntrinsicData{IntrinsicType::kPack2x16Unorm, ParamType::kF32, "metal::pack_float_to_unorm2x16"}, - IntrinsicData{semantic::Intrinsic::kPow, ParamType::kF32, "metal::pow"}, - IntrinsicData{semantic::Intrinsic::kReflect, ParamType::kF32, + IntrinsicData{IntrinsicType::kPow, ParamType::kF32, "metal::pow"}, + IntrinsicData{IntrinsicType::kReflect, ParamType::kF32, "metal::reflect"}, - IntrinsicData{semantic::Intrinsic::kReverseBits, ParamType::kU32, + IntrinsicData{IntrinsicType::kReverseBits, ParamType::kU32, "metal::reverse_bits"}, - IntrinsicData{semantic::Intrinsic::kRound, ParamType::kU32, - "metal::round"}, - IntrinsicData{semantic::Intrinsic::kSelect, ParamType::kF32, - "metal::select"}, - IntrinsicData{semantic::Intrinsic::kSign, ParamType::kF32, - "metal::sign"}, - IntrinsicData{semantic::Intrinsic::kSin, ParamType::kF32, "metal::sin"}, - IntrinsicData{semantic::Intrinsic::kSinh, ParamType::kF32, - "metal::sinh"}, - IntrinsicData{semantic::Intrinsic::kSmoothStep, ParamType::kF32, + IntrinsicData{IntrinsicType::kRound, ParamType::kU32, "metal::round"}, + IntrinsicData{IntrinsicType::kSelect, ParamType::kF32, "metal::select"}, + IntrinsicData{IntrinsicType::kSign, ParamType::kF32, "metal::sign"}, + IntrinsicData{IntrinsicType::kSin, ParamType::kF32, "metal::sin"}, + IntrinsicData{IntrinsicType::kSinh, ParamType::kF32, "metal::sinh"}, + IntrinsicData{IntrinsicType::kSmoothStep, ParamType::kF32, "metal::smoothstep"}, - IntrinsicData{semantic::Intrinsic::kSqrt, ParamType::kF32, - "metal::sqrt"}, - IntrinsicData{semantic::Intrinsic::kStep, ParamType::kF32, - "metal::step"}, - IntrinsicData{semantic::Intrinsic::kTan, ParamType::kF32, "metal::tan"}, - IntrinsicData{semantic::Intrinsic::kTanh, ParamType::kF32, - "metal::tanh"}, - IntrinsicData{semantic::Intrinsic::kTrunc, ParamType::kF32, - "metal::trunc"})); + IntrinsicData{IntrinsicType::kSqrt, ParamType::kF32, "metal::sqrt"}, + IntrinsicData{IntrinsicType::kStep, ParamType::kF32, "metal::step"}, + IntrinsicData{IntrinsicType::kTan, ParamType::kF32, "metal::tan"}, + IntrinsicData{IntrinsicType::kTanh, ParamType::kF32, "metal::tanh"}, + IntrinsicData{IntrinsicType::kTrunc, ParamType::kF32, "metal::trunc"})); TEST_F(MslGeneratorImplTest, Intrinsic_Call) { Global("param1", ast::StorageClass::kFunction, ty.vec2()); diff --git a/src/writer/spirv/builder.cc b/src/writer/spirv/builder.cc index a3513577aa..ce32ebb14e 100644 --- a/src/writer/spirv/builder.cc +++ b/src/writer/spirv/builder.cc @@ -87,6 +87,8 @@ namespace writer { namespace spirv { namespace { +using IntrinsicType = semantic::IntrinsicType; + const char kGLSLstd450[] = "GLSL.std.450"; uint32_t size_of(const InstructionList& instructions) { @@ -166,26 +168,25 @@ type::Matrix* GetNestedMatrixType(type::Type* type) { return type->As(); } -uint32_t intrinsic_to_glsl_method(type::Type* type, - semantic::Intrinsic intrinsic) { +uint32_t intrinsic_to_glsl_method(type::Type* type, IntrinsicType intrinsic) { switch (intrinsic) { - case semantic::Intrinsic::kAbs: + case IntrinsicType::kAbs: if (type->is_float_scalar_or_vector()) { return GLSLstd450FAbs; } else { return GLSLstd450SAbs; } - case semantic::Intrinsic::kAcos: + case IntrinsicType::kAcos: return GLSLstd450Acos; - case semantic::Intrinsic::kAsin: + case IntrinsicType::kAsin: return GLSLstd450Asin; - case semantic::Intrinsic::kAtan: + case IntrinsicType::kAtan: return GLSLstd450Atan; - case semantic::Intrinsic::kAtan2: + case IntrinsicType::kAtan2: return GLSLstd450Atan2; - case semantic::Intrinsic::kCeil: + case IntrinsicType::kCeil: return GLSLstd450Ceil; - case semantic::Intrinsic::kClamp: + case IntrinsicType::kClamp: if (type->is_float_scalar_or_vector()) { return GLSLstd450NClamp; } else if (type->is_unsigned_scalar_or_vector()) { @@ -193,41 +194,41 @@ uint32_t intrinsic_to_glsl_method(type::Type* type, } else { return GLSLstd450SClamp; } - case semantic::Intrinsic::kCos: + case IntrinsicType::kCos: return GLSLstd450Cos; - case semantic::Intrinsic::kCosh: + case IntrinsicType::kCosh: return GLSLstd450Cosh; - case semantic::Intrinsic::kCross: + case IntrinsicType::kCross: return GLSLstd450Cross; - case semantic::Intrinsic::kDeterminant: + case IntrinsicType::kDeterminant: return GLSLstd450Determinant; - case semantic::Intrinsic::kDistance: + case IntrinsicType::kDistance: return GLSLstd450Distance; - case semantic::Intrinsic::kExp: + case IntrinsicType::kExp: return GLSLstd450Exp; - case semantic::Intrinsic::kExp2: + case IntrinsicType::kExp2: return GLSLstd450Exp2; - case semantic::Intrinsic::kFaceForward: + case IntrinsicType::kFaceForward: return GLSLstd450FaceForward; - case semantic::Intrinsic::kFloor: + case IntrinsicType::kFloor: return GLSLstd450Floor; - case semantic::Intrinsic::kFma: + case IntrinsicType::kFma: return GLSLstd450Fma; - case semantic::Intrinsic::kFract: + case IntrinsicType::kFract: return GLSLstd450Fract; - case semantic::Intrinsic::kFrexp: + case IntrinsicType::kFrexp: return GLSLstd450Frexp; - case semantic::Intrinsic::kInverseSqrt: + case IntrinsicType::kInverseSqrt: return GLSLstd450InverseSqrt; - case semantic::Intrinsic::kLdexp: + case IntrinsicType::kLdexp: return GLSLstd450Ldexp; - case semantic::Intrinsic::kLength: + case IntrinsicType::kLength: return GLSLstd450Length; - case semantic::Intrinsic::kLog: + case IntrinsicType::kLog: return GLSLstd450Log; - case semantic::Intrinsic::kLog2: + case IntrinsicType::kLog2: return GLSLstd450Log2; - case semantic::Intrinsic::kMax: + case IntrinsicType::kMax: if (type->is_float_scalar_or_vector()) { return GLSLstd450NMax; } else if (type->is_unsigned_scalar_or_vector()) { @@ -235,7 +236,7 @@ uint32_t intrinsic_to_glsl_method(type::Type* type, } else { return GLSLstd450SMax; } - case semantic::Intrinsic::kMin: + case IntrinsicType::kMin: if (type->is_float_scalar_or_vector()) { return GLSLstd450NMin; } else if (type->is_unsigned_scalar_or_vector()) { @@ -243,45 +244,45 @@ uint32_t intrinsic_to_glsl_method(type::Type* type, } else { return GLSLstd450SMin; } - case semantic::Intrinsic::kMix: + case IntrinsicType::kMix: return GLSLstd450FMix; - case semantic::Intrinsic::kModf: + case IntrinsicType::kModf: return GLSLstd450Modf; - case semantic::Intrinsic::kNormalize: + case IntrinsicType::kNormalize: return GLSLstd450Normalize; - case semantic::Intrinsic::kPack4x8Snorm: + case IntrinsicType::kPack4x8Snorm: return GLSLstd450PackSnorm4x8; - case semantic::Intrinsic::kPack4x8Unorm: + case IntrinsicType::kPack4x8Unorm: return GLSLstd450PackUnorm4x8; - case semantic::Intrinsic::kPack2x16Snorm: + case IntrinsicType::kPack2x16Snorm: return GLSLstd450PackSnorm2x16; - case semantic::Intrinsic::kPack2x16Unorm: + case IntrinsicType::kPack2x16Unorm: return GLSLstd450PackUnorm2x16; - case semantic::Intrinsic::kPack2x16Float: + case IntrinsicType::kPack2x16Float: return GLSLstd450PackHalf2x16; - case semantic::Intrinsic::kPow: + case IntrinsicType::kPow: return GLSLstd450Pow; - case semantic::Intrinsic::kReflect: + case IntrinsicType::kReflect: return GLSLstd450Reflect; - case semantic::Intrinsic::kRound: + case IntrinsicType::kRound: return GLSLstd450Round; - case semantic::Intrinsic::kSign: + case IntrinsicType::kSign: return GLSLstd450FSign; - case semantic::Intrinsic::kSin: + case IntrinsicType::kSin: return GLSLstd450Sin; - case semantic::Intrinsic::kSinh: + case IntrinsicType::kSinh: return GLSLstd450Sinh; - case semantic::Intrinsic::kSmoothStep: + case IntrinsicType::kSmoothStep: return GLSLstd450SmoothStep; - case semantic::Intrinsic::kSqrt: + case IntrinsicType::kSqrt: return GLSLstd450Sqrt; - case semantic::Intrinsic::kStep: + case IntrinsicType::kStep: return GLSLstd450Step; - case semantic::Intrinsic::kTan: + case IntrinsicType::kTan: return GLSLstd450Tan; - case semantic::Intrinsic::kTanh: + case IntrinsicType::kTanh: return GLSLstd450Tanh; - case semantic::Intrinsic::kTrunc: + case IntrinsicType::kTrunc: return GLSLstd450Trunc; default: break; @@ -1897,11 +1898,11 @@ uint32_t Builder::GenerateIntrinsic(ast::IdentifierExpression* ident, OperandList params = {Operand::Int(result_type_id), result}; spv::Op op = spv::Op::OpNop; - if (intrinsic == semantic::Intrinsic::kAny) { + if (intrinsic == IntrinsicType::kAny) { op = spv::Op::OpAny; - } else if (intrinsic == semantic::Intrinsic::kAll) { + } else if (intrinsic == IntrinsicType::kAll) { op = spv::Op::OpAll; - } else if (intrinsic == semantic::Intrinsic::kArrayLength) { + } else if (intrinsic == IntrinsicType::kArrayLength) { if (call->params().empty()) { error_ = "missing param for runtime array length"; return 0; @@ -1934,35 +1935,35 @@ uint32_t Builder::GenerateIntrinsic(ast::IdentifierExpression* ident, return 0; } return result_id; - } else if (intrinsic == semantic::Intrinsic::kCountOneBits) { + } else if (intrinsic == IntrinsicType::kCountOneBits) { op = spv::Op::OpBitCount; - } else if (intrinsic == semantic::Intrinsic::kDot) { + } else if (intrinsic == IntrinsicType::kDot) { op = spv::Op::OpDot; - } else if (intrinsic == semantic::Intrinsic::kDpdx) { + } else if (intrinsic == IntrinsicType::kDpdx) { op = spv::Op::OpDPdx; - } else if (intrinsic == semantic::Intrinsic::kDpdxCoarse) { + } else if (intrinsic == IntrinsicType::kDpdxCoarse) { op = spv::Op::OpDPdxCoarse; - } else if (intrinsic == semantic::Intrinsic::kDpdxFine) { + } else if (intrinsic == IntrinsicType::kDpdxFine) { op = spv::Op::OpDPdxFine; - } else if (intrinsic == semantic::Intrinsic::kDpdy) { + } else if (intrinsic == IntrinsicType::kDpdy) { op = spv::Op::OpDPdy; - } else if (intrinsic == semantic::Intrinsic::kDpdyCoarse) { + } else if (intrinsic == IntrinsicType::kDpdyCoarse) { op = spv::Op::OpDPdyCoarse; - } else if (intrinsic == semantic::Intrinsic::kDpdyFine) { + } else if (intrinsic == IntrinsicType::kDpdyFine) { op = spv::Op::OpDPdyFine; - } else if (intrinsic == semantic::Intrinsic::kFwidth) { + } else if (intrinsic == IntrinsicType::kFwidth) { op = spv::Op::OpFwidth; - } else if (intrinsic == semantic::Intrinsic::kFwidthCoarse) { + } else if (intrinsic == IntrinsicType::kFwidthCoarse) { op = spv::Op::OpFwidthCoarse; - } else if (intrinsic == semantic::Intrinsic::kFwidthFine) { + } else if (intrinsic == IntrinsicType::kFwidthFine) { op = spv::Op::OpFwidthFine; - } else if (intrinsic == semantic::Intrinsic::kIsInf) { + } else if (intrinsic == IntrinsicType::kIsInf) { op = spv::Op::OpIsInf; - } else if (intrinsic == semantic::Intrinsic::kIsNan) { + } else if (intrinsic == IntrinsicType::kIsNan) { op = spv::Op::OpIsNan; - } else if (intrinsic == semantic::Intrinsic::kReverseBits) { + } else if (intrinsic == IntrinsicType::kReverseBits) { op = spv::Op::OpBitReverse; - } else if (intrinsic == semantic::Intrinsic::kSelect) { + } else if (intrinsic == IntrinsicType::kSelect) { op = spv::Op::OpSelect; } else { GenerateGLSLstd450Import(); @@ -2163,7 +2164,7 @@ bool Builder::GenerateTextureIntrinsic( }; switch (sem->intrinsic()) { - case semantic::Intrinsic::kTextureDimensions: { + case IntrinsicType::kTextureDimensions: { // Number of returned elements from OpImageQuerySize[Lod] may not match // those of textureDimensions(). // This might be due to an extra vector scalar describing the number of @@ -2219,7 +2220,7 @@ bool Builder::GenerateTextureIntrinsic( } break; } - case semantic::Intrinsic::kTextureNumLayers: { + case IntrinsicType::kTextureNumLayers: { uint32_t spirv_dims = 0; switch (texture_type->dim()) { default: @@ -2254,19 +2255,19 @@ bool Builder::GenerateTextureIntrinsic( } break; } - case semantic::Intrinsic::kTextureNumLevels: { + case IntrinsicType::kTextureNumLevels: { op = spv::Op::OpImageQueryLevels; append_result_type_and_id_to_spirv_params(); spirv_params.emplace_back(gen_param(pidx.texture)); break; } - case semantic::Intrinsic::kTextureNumSamples: { + case IntrinsicType::kTextureNumSamples: { op = spv::Op::OpImageQuerySamples; append_result_type_and_id_to_spirv_params(); spirv_params.emplace_back(gen_param(pidx.texture)); break; } - case semantic::Intrinsic::kTextureLoad: { + case IntrinsicType::kTextureLoad: { op = texture_type->Is() ? spv::Op::OpImageRead : spv::Op::OpImageFetch; append_result_type_and_id_to_spirv_params_for_read(); @@ -2287,7 +2288,7 @@ bool Builder::GenerateTextureIntrinsic( break; } - case semantic::Intrinsic::kTextureStore: { + case IntrinsicType::kTextureStore: { op = spv::Op::OpImageWrite; spirv_params.emplace_back(gen_param(pidx.texture)); if (!append_coords_to_spirv_params()) { @@ -2296,7 +2297,7 @@ bool Builder::GenerateTextureIntrinsic( spirv_params.emplace_back(gen_param(pidx.value)); break; } - case semantic::Intrinsic::kTextureSample: { + case IntrinsicType::kTextureSample: { op = spv::Op::OpImageSampleImplicitLod; append_result_type_and_id_to_spirv_params_for_read(); if (!append_image_and_coords_to_spirv_params()) { @@ -2304,7 +2305,7 @@ bool Builder::GenerateTextureIntrinsic( } break; } - case semantic::Intrinsic::kTextureSampleBias: { + case IntrinsicType::kTextureSampleBias: { op = spv::Op::OpImageSampleImplicitLod; append_result_type_and_id_to_spirv_params_for_read(); if (!append_image_and_coords_to_spirv_params()) { @@ -2315,7 +2316,7 @@ bool Builder::GenerateTextureIntrinsic( ImageOperand{SpvImageOperandsBiasMask, gen_param(pidx.bias)}); break; } - case semantic::Intrinsic::kTextureSampleLevel: { + case IntrinsicType::kTextureSampleLevel: { op = spv::Op::OpImageSampleExplicitLod; append_result_type_and_id_to_spirv_params_for_read(); if (!append_image_and_coords_to_spirv_params()) { @@ -2339,7 +2340,7 @@ bool Builder::GenerateTextureIntrinsic( image_operands.emplace_back(ImageOperand{SpvImageOperandsLodMask, level}); break; } - case semantic::Intrinsic::kTextureSampleGrad: { + case IntrinsicType::kTextureSampleGrad: { op = spv::Op::OpImageSampleExplicitLod; append_result_type_and_id_to_spirv_params_for_read(); if (!append_image_and_coords_to_spirv_params()) { @@ -2353,7 +2354,7 @@ bool Builder::GenerateTextureIntrinsic( ImageOperand{SpvImageOperandsGradMask, gen_param(pidx.ddy)}); break; } - case semantic::Intrinsic::kTextureSampleCompare: { + case IntrinsicType::kTextureSampleCompare: { op = spv::Op::OpImageSampleDrefExplicitLod; append_result_type_and_id_to_spirv_params(); if (!append_image_and_coords_to_spirv_params()) {