encounter
/
dawn-cmake
mirror of https://github.com/encounter/dawn-cmake.git
1
0
Fork 0
Code Issues Packages Projects Releases Wiki Activity

tint: Implement const-eval of frexp() 

Also add abstract overloads.

Bug: tint:1581
Fixed: tint:1768
Change-Id: Icda465e0cfe960b77823c2135f0cfe8f82ed394f
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/111441
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Reviewed-by: Dan Sinclair <dsinclair@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
This commit is contained in:
Ben Clayton 2022-11-23 18:21:38 +00:00 committed by Dawn LUCI CQ
parent d743778ed5
commit 69c2c34326
218 changed files with 6625 additions and 719 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

4
src/tint/intrinsics.def
View File

@ -477,8 +477,8 @@ fn fma<T: f32_f16>(T, T, T) -> T
fn fma<N: num, T: f32_f16>(vec<N, T>, vec<N, T>, vec<N, T>) -> vec<N, T>
fn fract<T: f32_f16>(T) -> T
fn fract<N: num, T: f32_f16>(vec<N, T>) -> vec<N, T>
fn frexp<T: f32_f16>(T) -> __frexp_result<T>
fn frexp<N: num, T: f32_f16>(vec<N, T>) -> __frexp_result_vec<N, T>
@const fn frexp<T: fa_f32_f16>(T) -> __frexp_result<T>
@const fn frexp<N: num, T: fa_f32_f16>(vec<N, T>) -> __frexp_result_vec<N, T>
@stage("fragment") fn fwidth(f32) -> f32
@stage("fragment") fn fwidth<N: num>(vec<N, f32>) -> vec<N, f32>
@stage("fragment") fn fwidthCoarse(f32) -> f32

50
src/tint/resolver/builtin_test.cc
View File

@ -1079,54 +1079,8 @@ TEST_F(ResolverBuiltinFloatTest, Frexp_Error_FirstParamInt) {
R"(error: no matching call to frexp(i32, ptr<workgroup, i32, read_write>)
2 candidate functions:
frexp(T) -> __frexp_result_T where: T is f32 or f16
frexp(vecN<T>) -> __frexp_result_vecN_T where: T is f32 or f16
)");
}
TEST_F(ResolverBuiltinFloatTest, Frexp_Error_SecondParamFloatPtr) {
GlobalVar("v", ty.f32(), ast::AddressSpace::kWorkgroup);
auto* call = Call("frexp", 1_f, AddressOf("v"));
WrapInFunction(call);
EXPECT_FALSE(r()->Resolve());
EXPECT_EQ(r()->error(),
R"(error: no matching call to frexp(f32, ptr<workgroup, f32, read_write>)
2 candidate functions:
frexp(T) -> __frexp_result_T where: T is f32 or f16
frexp(vecN<T>) -> __frexp_result_vecN_T where: T is f32 or f16
)");
}
TEST_F(ResolverBuiltinFloatTest, Frexp_Error_SecondParamNotAPointer) {
auto* call = Call("frexp", 1_f, 1_i);
WrapInFunction(call);
EXPECT_FALSE(r()->Resolve());
EXPECT_EQ(r()->error(), R"(error: no matching call to frexp(f32, i32)
2 candidate functions:
frexp(T) -> __frexp_result_T where: T is f32 or f16
frexp(vecN<T>) -> __frexp_result_vecN_T where: T is f32 or f16
)");
}
TEST_F(ResolverBuiltinFloatTest, Frexp_Error_VectorSizesDontMatch) {
GlobalVar("v", ty.vec4<i32>(), ast::AddressSpace::kWorkgroup);
auto* call = Call("frexp", vec2<f32>(1_f, 2_f), AddressOf("v"));
WrapInFunction(call);
EXPECT_FALSE(r()->Resolve());
EXPECT_EQ(r()->error(),
R"(error: no matching call to frexp(vec2<f32>, ptr<workgroup, vec4<i32>, read_write>)
2 candidate functions:
frexp(T) -> __frexp_result_T where: T is f32 or f16
frexp(vecN<T>) -> __frexp_result_vecN_T where: T is f32 or f16
frexp(T) -> __frexp_result_T where: T is abstract-float, f32 or f16
frexp(vecN<T>) -> __frexp_result_vecN_T where: T is abstract-float, f32 or f16
)");
}

73
src/tint/resolver/const_eval.cc
View File

@ -2219,6 +2219,79 @@ ConstEval::Result ConstEval::floor(const sem::Type* ty,
return TransformElements(builder, ty, transform, args[0]);
}
ConstEval::Result ConstEval::frexp(const sem::Type* ty,
utils::VectorRef<const sem::Constant*> args,
const Source& source) {
auto* arg = args[0];
struct FractExp {
ImplResult fract;
ImplResult exp;
};
auto scalar = [&](const sem::Constant* s) {
int exp = 0;
double fract = std::frexp(s->As<AFloat>(), &exp);
return Switch(
s->Type(),
[&](const sem::F32*) {
return FractExp{
CreateElement(builder, source, builder.create<sem::F32>(), f32(fract)),
CreateElement(builder, source, builder.create<sem::I32>(), i32(exp)),
};
},
[&](const sem::F16*) {
return FractExp{
CreateElement(builder, source, builder.create<sem::F16>(), f16(fract)),
CreateElement(builder, source, builder.create<sem::I32>(), i32(exp)),
};
},
[&](const sem::AbstractFloat*) {
return FractExp{
CreateElement(builder, source, builder.create<sem::AbstractFloat>(),
AFloat(fract)),
CreateElement(builder, source, builder.create<sem::AbstractInt>(), AInt(exp)),
};
},
[&](Default) {
TINT_ICE(Resolver, builder.Diagnostics())
<< "unhandled element type for frexp() const-eval: "
<< builder.FriendlyName(s->Type());
return FractExp{utils::Failure, utils::Failure};
});
};
if (auto* vec = arg->Type()->As<sem::Vector>()) {
utils::Vector<const sem::Constant*, 4> fract_els;
utils::Vector<const sem::Constant*, 4> exp_els;
for (uint32_t i = 0; i < vec->Width(); i++) {
auto fe = scalar(arg->Index(i));
if (!fe.fract || !fe.exp) {
return utils::Failure;
}
fract_els.Push(fe.fract.Get());
exp_els.Push(fe.exp.Get());
}
auto fract_ty = builder.create<sem::Vector>(fract_els[0]->Type(), vec->Width());
auto exp_ty = builder.create<sem::Vector>(exp_els[0]->Type(), vec->Width());
return CreateComposite(builder, ty,
utils::Vector<const sem::Constant*, 2>{
CreateComposite(builder, fract_ty, std::move(fract_els)),
CreateComposite(builder, exp_ty, std::move(exp_els)),
});
} else {
auto fe = scalar(arg);
if (!fe.fract || !fe.exp) {
return utils::Failure;
}
return CreateComposite(builder, ty,
utils::Vector<const sem::Constant*, 2>{
fe.fract.Get(),
fe.exp.Get(),
});
}
}
ConstEval::Result ConstEval::insertBits(const sem::Type* ty,
utils::VectorRef<const sem::Constant*> args,
const Source& source) {

9
src/tint/resolver/const_eval.h
View File

@ -610,6 +610,15 @@ class ConstEval {
utils::VectorRef<const sem::Constant*> args,
const Source& source);
/// frexp builtin
/// @param ty the expression type
/// @param args the input arguments
/// @param source the source location
/// @return the result value, or null if the value cannot be calculated
Result frexp(const sem::Type* ty,
utils::VectorRef<const sem::Constant*> args,
const Source& source);
/// insertBits builtin
/// @param ty the expression type
/// @param args the input arguments

56
src/tint/resolver/const_eval_builtin_test.cc
View File

@ -968,6 +968,62 @@ INSTANTIATE_TEST_SUITE_P( //
testing::ValuesIn(Concat(FloorCases<AFloat>(), //
FloorCases<f32>(),
FloorCases<f16>()))));
template <typename T>
std::vector<Case> FrexpCases() {
using F = T; // fract type
using E = std::conditional_t<std::is_same_v<T, AFloat>, AInt, i32>; // exp type
auto cases = std::vector<Case>{
// Scalar tests
// in fract exp
C({T(-3.5)}, {F(-0.875), E(2)}), //
C({T(-3.0)}, {F(-0.750), E(2)}), //
C({T(-2.5)}, {F(-0.625), E(2)}), //
C({T(-2.0)}, {F(-0.500), E(2)}), //
C({T(-1.5)}, {F(-0.750), E(1)}), //
C({T(-1.0)}, {F(-0.500), E(1)}), //
C({T(+0.0)}, {F(+0.000), E(0)}), //
C({T(+1.0)}, {F(+0.500), E(1)}), //
C({T(+1.5)}, {F(+0.750), E(1)}), //
C({T(+2.0)}, {F(+0.500), E(2)}), //
C({T(+2.5)}, {F(+0.625), E(2)}), //
C({T(+3.0)}, {F(+0.750), E(2)}), //
C({T(+3.5)}, {F(+0.875), E(2)}), //
// Vector tests
// in fract exp
C({Vec(T(-2.5), T(+1.0))}, {Vec(F(-0.625), F(+0.500)), Vec(E(2), E(1))}),
C({Vec(T(+3.5), T(-2.5))}, {Vec(F(+0.875), F(-0.625)), Vec(E(2), E(2))}),
};
ConcatIntoIf<std::is_same_v<T, f16>>(cases, std::vector<Case>{
C({T::Highest()}, {F(0x0.ffep0), E(16)}), //
C({T::Lowest()}, {F(-0x0.ffep0), E(16)}), //
C({T::Smallest()}, {F(0.5), E(-13)}), //
});
ConcatIntoIf<std::is_same_v<T, f32>>(cases,
std::vector<Case>{
C({T::Highest()}, {F(0x0.ffffffp0), E(128)}), //
C({T::Lowest()}, {F(-0x0.ffffffp0), E(128)}), //
C({T::Smallest()}, {F(0.5), E(-125)}), //
});
ConcatIntoIf<std::is_same_v<T, AFloat>>(
cases, std::vector<Case>{
C({T::Highest()}, {F(0x0.fffffffffffff8p0), E(1024)}), //
C({T::Lowest()}, {F(-0x0.fffffffffffff8p0), E(1024)}), //
C({T::Smallest()}, {F(0.5), E(-1021)}), //
});
return cases;
}
INSTANTIATE_TEST_SUITE_P( //
Frexp,
ResolverConstEvalBuiltinTest,
testing::Combine(testing::Values(sem::BuiltinType::kFrexp),
testing::ValuesIn(Concat(FrexpCases<AFloat>(), //
FrexpCases<f32>(), //
FrexpCases<f16>()))));
template <typename T>
std::vector<Case> InsertBitsCases() {

78
src/tint/resolver/intrinsic_table.cc
View File

@ -855,6 +855,7 @@ const sem::Struct* build_modf_result(MatchState& state, const sem::Type* el) {
return nullptr;
});
}
const sem::Struct* build_modf_result_vec(MatchState& state, Number& n, const sem::Type* el) {
auto prefix = "__modf_result_vec" + std::to_string(n.Value());
auto build_f32 = [&] {
@ -883,27 +884,70 @@ const sem::Struct* build_modf_result_vec(MatchState& state, Number& n, const sem
return nullptr;
});
}
const sem::Struct* build_frexp_result(MatchState& state, const sem::Type* el) {
std::string display_name;
if (el->Is<sem::F16>()) {
display_name = "__frexp_result_f16";
} else {
display_name = "__frexp_result";
}
auto* i32 = state.builder.create<sem::I32>();
return build_struct(state.builder, display_name, {{"fract", el}, {"exp", i32}});
auto build_f32 = [&] {
auto* f = state.builder.create<sem::F32>();
auto* i = state.builder.create<sem::I32>();
return build_struct(state.builder, "__frexp_result", {{"fract", f}, {"exp", i}});
};
auto build_f16 = [&] {
auto* f = state.builder.create<sem::F16>();
auto* i = state.builder.create<sem::I32>();
return build_struct(state.builder, "__frexp_result_f16", {{"fract", f}, {"exp", i}});
};
return Switch(
el, //
[&](const sem::F32*) { return build_f32(); }, //
[&](const sem::F16*) { return build_f16(); }, //
[&](const sem::AbstractFloat*) {
auto* i = state.builder.create<sem::AbstractInt>();
auto* abstract =
build_struct(state.builder, "__frexp_result_abstract", {{"fract", el}, {"exp", i}});
abstract->SetConcreteTypes(utils::Vector{build_f32(), build_f16()});
return abstract;
},
[&](Default) {
TINT_ICE(Resolver, state.builder.Diagnostics())
<< "unhandled frexp type: " << state.builder.FriendlyName(el);
return nullptr;
});
}
const sem::Struct* build_frexp_result_vec(MatchState& state, Number& n, const sem::Type* el) {
std::string display_name;
if (el->Is<sem::F16>()) {
display_name = "__frexp_result_vec" + std::to_string(n.Value()) + "_f16";
} else {
display_name = "__frexp_result_vec" + std::to_string(n.Value());
}
auto* vec = state.builder.create<sem::Vector>(el, n.Value());
auto* vec_i32 = state.builder.create<sem::Vector>(state.builder.create<sem::I32>(), n.Value());
return build_struct(state.builder, display_name, {{"fract", vec}, {"exp", vec_i32}});
auto prefix = "__frexp_result_vec" + std::to_string(n.Value());
auto build_f32 = [&] {
auto* f = state.builder.create<sem::Vector>(state.builder.create<sem::F32>(), n.Value());
auto* e = state.builder.create<sem::Vector>(state.builder.create<sem::I32>(), n.Value());
return build_struct(state.builder, prefix, {{"fract", f}, {"exp", e}});
};
auto build_f16 = [&] {
auto* f = state.builder.create<sem::Vector>(state.builder.create<sem::F16>(), n.Value());
auto* e = state.builder.create<sem::Vector>(state.builder.create<sem::I32>(), n.Value());
return build_struct(state.builder, prefix + "_f16", {{"fract", f}, {"exp", e}});
};
return Switch(
el, //
[&](const sem::F32*) { return build_f32(); }, //
[&](const sem::F16*) { return build_f16(); }, //
[&](const sem::AbstractFloat*) {
auto* f = state.builder.create<sem::Vector>(el, n.Value());
auto* e = state.builder.create<sem::Vector>(state.builder.create<sem::AbstractInt>(),
n.Value());
auto* abstract =
build_struct(state.builder, prefix + "_abstract", {{"fract", f}, {"exp", e}});
abstract->SetConcreteTypes(utils::Vector{build_f32(), build_f16()});
return abstract;
},
[&](Default) {
TINT_ICE(Resolver, state.builder.Diagnostics())
<< "unhandled frexp type: " << state.builder.FriendlyName(el);
return nullptr;
});
}
const sem::Struct* build_atomic_compare_exchange_result(MatchState& state, const sem::Type* ty) {
return build_struct(
state.builder,

12
src/tint/resolver/intrinsic_table.inl
View File

@ -12558,24 +12558,24 @@ constexpr OverloadInfo kOverloads[] = {
/* num parameters */ 1,
/* num template types */ 1,
/* num template numbers */ 0,
/* template types */ &kTemplateTypes[26],
/* template types */ &kTemplateTypes[23],
/* template numbers */ &kTemplateNumbers[10],
/* parameters */ &kParameters[862],
/* return matcher indices */ &kMatcherIndices[104],
/* flags */ OverloadFlags(OverloadFlag::kIsBuiltin, OverloadFlag::kSupportsVertexPipeline, OverloadFlag::kSupportsFragmentPipeline, OverloadFlag::kSupportsComputePipeline),
/* const eval */ nullptr,
/* const eval */ &ConstEval::frexp,
},
{
/* [354] */
/* num parameters */ 1,
/* num template types */ 1,
/* num template numbers */ 1,
/* template types */ &kTemplateTypes[26],
/* template types */ &kTemplateTypes[23],
/* template numbers */ &kTemplateNumbers[4],
/* parameters */ &kParameters[863],
/* return matcher indices */ &kMatcherIndices[39],
/* flags */ OverloadFlags(OverloadFlag::kIsBuiltin, OverloadFlag::kSupportsVertexPipeline, OverloadFlag::kSupportsFragmentPipeline, OverloadFlag::kSupportsComputePipeline),
/* const eval */ nullptr,
/* const eval */ &ConstEval::frexp,
},
{
/* [355] */
@ -14259,8 +14259,8 @@ constexpr IntrinsicInfo kBuiltins[] = {
},
{
/* [40] */
/* fn frexp<T : f32_f16>(T) -> __frexp_result<T> */
/* fn frexp<N : num, T : f32_f16>(vec<N, T>) -> __frexp_result_vec<N, T> */
/* fn frexp<T : fa_f32_f16>(T) -> __frexp_result<T> */
/* fn frexp<N : num, T : fa_f32_f16>(vec<N, T>) -> __frexp_result_vec<N, T> */
/* num overloads */ 2,
/* overloads */ &kOverloads[353],
},

88
src/tint/resolver/materialize_test.cc
View File

@ -1320,6 +1320,94 @@ TEST_F(MaterializeAbstractStructure, Modf_Vector_ExplicitType) {
abstract_str->Members()[0]->Type()->As<sem::Vector>()->type()->Is<sem::AbstractFloat>());
}
TEST_F(MaterializeAbstractStructure, Frexp_Scalar_DefaultType) {
// var v = frexp(1);
auto* call = Call("frexp", 1_a);
WrapInFunction(Decl(Var("v", call)));
ASSERT_TRUE(r()->Resolve()) << r()->error();
auto* sem = Sem().Get(call);
ASSERT_TRUE(sem->Is<sem::Materialize>());
auto* materialize = sem->As<sem::Materialize>();
ASSERT_TRUE(materialize->Type()->Is<sem::Struct>());
auto* concrete_str = materialize->Type()->As<sem::Struct>();
ASSERT_TRUE(concrete_str->Members()[0]->Type()->Is<sem::F32>());
ASSERT_TRUE(concrete_str->Members()[1]->Type()->Is<sem::I32>());
ASSERT_TRUE(materialize->Expr()->Type()->Is<sem::Struct>());
auto* abstract_str = materialize->Expr()->Type()->As<sem::Struct>();
ASSERT_TRUE(abstract_str->Members()[0]->Type()->Is<sem::AbstractFloat>());
ASSERT_TRUE(abstract_str->Members()[1]->Type()->Is<sem::AbstractInt>());
}
TEST_F(MaterializeAbstractStructure, Frexp_Vector_DefaultType) {
// var v = frexp(vec2(1));
auto* call = Call("frexp", Construct(ty.vec2(nullptr), 1_a));
WrapInFunction(Decl(Var("v", call)));
ASSERT_TRUE(r()->Resolve()) << r()->error();
auto* sem = Sem().Get(call);
ASSERT_TRUE(sem->Is<sem::Materialize>());
auto* materialize = sem->As<sem::Materialize>();
ASSERT_TRUE(materialize->Type()->Is<sem::Struct>());
auto* concrete_str = materialize->Type()->As<sem::Struct>();
ASSERT_TRUE(concrete_str->Members()[0]->Type()->Is<sem::Vector>());
ASSERT_TRUE(concrete_str->Members()[1]->Type()->Is<sem::Vector>());
ASSERT_TRUE(concrete_str->Members()[0]->Type()->As<sem::Vector>()->type()->Is<sem::F32>());
ASSERT_TRUE(concrete_str->Members()[1]->Type()->As<sem::Vector>()->type()->Is<sem::I32>());
ASSERT_TRUE(materialize->Expr()->Type()->Is<sem::Struct>());
auto* abstract_str = materialize->Expr()->Type()->As<sem::Struct>();
ASSERT_TRUE(abstract_str->Members()[0]->Type()->Is<sem::Vector>());
ASSERT_TRUE(
abstract_str->Members()[0]->Type()->As<sem::Vector>()->type()->Is<sem::AbstractFloat>());
ASSERT_TRUE(
abstract_str->Members()[1]->Type()->As<sem::Vector>()->type()->Is<sem::AbstractInt>());
}
TEST_F(MaterializeAbstractStructure, Frexp_Scalar_ExplicitType) {
// var v = frexp(1_h); // v is __frexp_result_f16
// v = frexp(1); // __frexp_result_f16 <- __frexp_result_abstract
Enable(ast::Extension::kF16);
auto* call = Call("frexp", 1_a);
WrapInFunction(Decl(Var("v", Call("frexp", 1_h))), //
Assign("v", call));
ASSERT_TRUE(r()->Resolve()) << r()->error();
auto* sem = Sem().Get(call);
ASSERT_TRUE(sem->Is<sem::Materialize>());
auto* materialize = sem->As<sem::Materialize>();
ASSERT_TRUE(materialize->Type()->Is<sem::Struct>());
auto* concrete_str = materialize->Type()->As<sem::Struct>();
ASSERT_TRUE(concrete_str->Members()[0]->Type()->Is<sem::F16>());
ASSERT_TRUE(concrete_str->Members()[1]->Type()->Is<sem::I32>());
ASSERT_TRUE(materialize->Expr()->Type()->Is<sem::Struct>());
auto* abstract_str = materialize->Expr()->Type()->As<sem::Struct>();
ASSERT_TRUE(abstract_str->Members()[0]->Type()->Is<sem::AbstractFloat>());
ASSERT_TRUE(abstract_str->Members()[1]->Type()->Is<sem::AbstractInt>());
}
TEST_F(MaterializeAbstractStructure, Frexp_Vector_ExplicitType) {
// var v = frexp(vec2(1_h)); // v is __frexp_result_vec2_f16
// v = frexp(vec2(1)); // __frexp_result_vec2_f16 <- __frexp_result_vec2_abstract
Enable(ast::Extension::kF16);
auto* call = Call("frexp", Construct(ty.vec2(nullptr), 1_a));
WrapInFunction(Decl(Var("v", Call("frexp", Construct(ty.vec2(nullptr), 1_h)))),
Assign("v", call));
ASSERT_TRUE(r()->Resolve()) << r()->error();
auto* sem = Sem().Get(call);
ASSERT_TRUE(sem->Is<sem::Materialize>());
auto* materialize = sem->As<sem::Materialize>();
ASSERT_TRUE(materialize->Type()->Is<sem::Struct>());
auto* concrete_str = materialize->Type()->As<sem::Struct>();
ASSERT_TRUE(concrete_str->Members()[0]->Type()->Is<sem::Vector>());
ASSERT_TRUE(concrete_str->Members()[1]->Type()->Is<sem::Vector>());
ASSERT_TRUE(concrete_str->Members()[0]->Type()->As<sem::Vector>()->type()->Is<sem::F16>());
ASSERT_TRUE(concrete_str->Members()[1]->Type()->As<sem::Vector>()->type()->Is<sem::I32>());
ASSERT_TRUE(materialize->Expr()->Type()->Is<sem::Struct>());
auto* abstract_str = materialize->Expr()->Type()->As<sem::Struct>();
ASSERT_TRUE(abstract_str->Members()[0]->Type()->Is<sem::Vector>());
ASSERT_TRUE(
abstract_str->Members()[0]->Type()->As<sem::Vector>()->type()->Is<sem::AbstractFloat>());
ASSERT_TRUE(
abstract_str->Members()[1]->Type()->As<sem::Vector>()->type()->Is<sem::AbstractInt>());
}
} // namespace materialize_abstract_structure
} // namespace

160
src/tint/writer/glsl/generator_impl_builtin_test.cc
View File

@ -668,9 +668,9 @@ void main() {
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Frexp_Scalar_f32) {
auto* call = Call("frexp", 1_f);
WrapInFunction(CallStmt(call));
TEST_F(GlslGeneratorImplTest_Builtin, Runtime_Frexp_Scalar_f32) {
WrapInFunction(Var("f", Expr(1_f)), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -690,7 +690,8 @@ frexp_result tint_frexp(float param_0) {
void test_function() {
tint_frexp(1.0f);
float f = 1.0f;
frexp_result v = tint_frexp(f);
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
@ -701,11 +702,11 @@ void main() {
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Frexp_Scalar_f16) {
TEST_F(GlslGeneratorImplTest_Builtin, Runtime_Frexp_Scalar_f16) {
Enable(ast::Extension::kF16);
auto* call = Call("frexp", 1_h);
WrapInFunction(CallStmt(call));
WrapInFunction(Var("f", Expr(1_h)), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -726,7 +727,8 @@ frexp_result_f16 tint_frexp(float16_t param_0) {
void test_function() {
tint_frexp(1.0hf);
float16_t f = 1.0hf;
frexp_result_f16 v = tint_frexp(f);
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
@ -737,9 +739,9 @@ void main() {
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Frexp_Vector_f32) {
auto* call = Call("frexp", vec3<f32>());
WrapInFunction(CallStmt(call));
TEST_F(GlslGeneratorImplTest_Builtin, Runtime_Frexp_Vector_f32) {
WrapInFunction(Var("f", Expr(vec3<f32>())), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -759,7 +761,8 @@ frexp_result_vec3 tint_frexp(vec3 param_0) {
void test_function() {
tint_frexp(vec3(0.0f));
vec3 f = vec3(0.0f);
frexp_result_vec3 v = tint_frexp(f);
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
@ -770,11 +773,11 @@ void main() {
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Frexp_Vector_f16) {
TEST_F(GlslGeneratorImplTest_Builtin, Runtime_Frexp_Vector_f16) {
Enable(ast::Extension::kF16);
auto* call = Call("frexp", vec3<f16>());
WrapInFunction(CallStmt(call));
WrapInFunction(Var("f", Expr(vec3<f16>())), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -795,7 +798,118 @@ frexp_result_vec3_f16 tint_frexp(f16vec3 param_0) {
void test_function() {
tint_frexp(f16vec3(0.0hf));
f16vec3 f = f16vec3(0.0hf);
frexp_result_vec3_f16 v = tint_frexp(f);
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
test_function();
return;
}
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Const_Frexp_Scalar_f32) {
WrapInFunction(Decl(Let("v", Call("frexp", 1_f))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#version 310 es
struct frexp_result {
float fract;
int exp;
};
void test_function() {
frexp_result v = frexp_result(0.5f, 1);
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
test_function();
return;
}
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Const_Frexp_Scalar_f16) {
Enable(ast::Extension::kF16);
WrapInFunction(Decl(Let("v", Call("frexp", 1_h))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#version 310 es
#extension GL_AMD_gpu_shader_half_float : require
struct frexp_result_f16 {
float16_t fract;
int exp;
};
void test_function() {
frexp_result_f16 v = frexp_result_f16(0.5hf, 1);
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
test_function();
return;
}
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Const_Frexp_Vector_f32) {
WrapInFunction(Decl(Let("v", Call("frexp", vec3<f32>()))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#version 310 es
struct frexp_result_vec3 {
vec3 fract;
ivec3 exp;
};
void test_function() {
frexp_result_vec3 v = frexp_result_vec3(vec3(0.0f), ivec3(0));
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
test_function();
return;
}
)");
}
TEST_F(GlslGeneratorImplTest_Builtin, Const_Frexp_Vector_f16) {
Enable(ast::Extension::kF16);
WrapInFunction(Decl(Let("v", Call("frexp", vec3<f16>()))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#version 310 es
#extension GL_AMD_gpu_shader_half_float : require
struct frexp_result_vec3_f16 {
f16vec3 fract;
ivec3 exp;
};
void test_function() {
frexp_result_vec3_f16 v = frexp_result_vec3_f16(f16vec3(0.0hf), ivec3(0));
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
@ -815,20 +929,8 @@ TEST_F(GlslGeneratorImplTest_Builtin, Frexp_Sig_Deprecation) {
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#version 310 es
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void test_function() {
float tint_symbol = tint_frexp(1.0f).fract;
float tint_symbol = 0.5f;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;

157
src/tint/writer/hlsl/generator_impl_builtin_test.cc
View File

@ -572,7 +572,7 @@ TEST_F(HlslGeneratorImplTest_Builtin, NonInitializer_Modf_Vector_f32) {
Decl(Var("v", Call("modf", vec3<f32>(1.5_f, 2.5_f, 3.5_f)))),
// Now assign 'v' again with another modf call.
// This requires generating a temporary variable for the struct initializer.
Assign("v", Call("modf", vec3<f32>(4.5_f, 5.5_f, 6.5_f))));
Assign("v", Call("modf", vec3<f32>(4.5_a, 5.5_a, 6.5_a))));
GeneratorImpl& gen = SanitizeAndBuild();
@ -591,9 +591,9 @@ void test_function() {
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, Frexp_Scalar_f32) {
auto* call = Call("frexp", 1_f);
WrapInFunction(CallStmt(call));
TEST_F(HlslGeneratorImplTest_Builtin, Runtime_Frexp_Scalar_f32) {
WrapInFunction(Var("f", Expr(1_f)), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -611,17 +611,18 @@ frexp_result tint_frexp(float param_0) {
[numthreads(1, 1, 1)]
void test_function() {
tint_frexp(1.0f);
float f = 1.0f;
frexp_result v = tint_frexp(f);
return;
}
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, Frexp_Scalar_f16) {
TEST_F(HlslGeneratorImplTest_Builtin, Runtime_Frexp_Scalar_f16) {
Enable(ast::Extension::kF16);
auto* call = Call("frexp", 1_h);
WrapInFunction(CallStmt(call));
WrapInFunction(Var("f", Expr(1_h)), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -639,15 +640,16 @@ frexp_result_f16 tint_frexp(float16_t param_0) {
[numthreads(1, 1, 1)]
void test_function() {
tint_frexp(float16_t(1.0h));
float16_t f = float16_t(1.0h);
frexp_result_f16 v = tint_frexp(f);
return;
}
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, Frexp_Vector_f32) {
auto* call = Call("frexp", vec3<f32>());
WrapInFunction(CallStmt(call));
TEST_F(HlslGeneratorImplTest_Builtin, Runtime_Frexp_Vector_f32) {
WrapInFunction(Var("f", Expr(vec3<f32>())), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -665,17 +667,18 @@ frexp_result_vec3 tint_frexp(float3 param_0) {
[numthreads(1, 1, 1)]
void test_function() {
tint_frexp((0.0f).xxx);
float3 f = (0.0f).xxx;
frexp_result_vec3 v = tint_frexp(f);
return;
}
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, Frexp_Vector_f16) {
TEST_F(HlslGeneratorImplTest_Builtin, Runtime_Frexp_Vector_f16) {
Enable(ast::Extension::kF16);
auto* call = Call("frexp", vec3<f16>());
WrapInFunction(CallStmt(call));
WrapInFunction(Var("f", Expr(vec3<f16>())), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -693,15 +696,15 @@ frexp_result_vec3_f16 tint_frexp(vector<float16_t, 3> param_0) {
[numthreads(1, 1, 1)]
void test_function() {
tint_frexp((float16_t(0.0h)).xxx);
vector<float16_t, 3> f = (float16_t(0.0h)).xxx;
frexp_result_vec3_f16 v = tint_frexp(f);
return;
}
)");
}
// TODO(crbug.com/tint/1757): Remove once deprecation period for `frexp().sig` is over
TEST_F(HlslGeneratorImplTest_Builtin, Frexp_Sig_Deprecation) {
WrapInFunction(MemberAccessor(Call("frexp", 1_f), "sig"));
TEST_F(HlslGeneratorImplTest_Builtin, Const_Frexp_Scalar_f32) {
WrapInFunction(Decl(Let("v", Call("frexp", 1_f))));
GeneratorImpl& gen = SanitizeAndBuild();
@ -710,16 +713,114 @@ TEST_F(HlslGeneratorImplTest_Builtin, Frexp_Sig_Deprecation) {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void test_function() {
const float tint_symbol = tint_frexp(1.0f).fract;
const frexp_result v = {0.5f, 1};
return;
}
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, Const_Frexp_Scalar_f16) {
Enable(ast::Extension::kF16);
WrapInFunction(Decl(Let("v", Call("frexp", 1_h))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(struct frexp_result_f16 {
float16_t fract;
int exp;
};
[numthreads(1, 1, 1)]
void test_function() {
const frexp_result_f16 v = {float16_t(0.5h), 1};
return;
}
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, Const_Frexp_Vector_f32) {
WrapInFunction(Decl(Let("v", Call("frexp", vec3<f32>()))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(struct frexp_result_vec3 {
float3 fract;
int3 exp;
};
[numthreads(1, 1, 1)]
void test_function() {
const frexp_result_vec3 v = (frexp_result_vec3)0;
return;
}
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, Const_Frexp_Vector_f16) {
Enable(ast::Extension::kF16);
WrapInFunction(Decl(Let("v", Call("frexp", vec3<f16>()))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(struct frexp_result_vec3_f16 {
vector<float16_t, 3> fract;
int3 exp;
};
[numthreads(1, 1, 1)]
void test_function() {
const frexp_result_vec3_f16 v = (frexp_result_vec3_f16)0;
return;
}
)");
}
TEST_F(HlslGeneratorImplTest_Builtin, NonInitializer_Frexp_Vector_f32) {
WrapInFunction(
// Declare a variable with the result of a frexp call.
// This is required to infer the 'var' type.
Decl(Var("v", Call("frexp", vec3<f32>(1.5_f, 2.5_f, 3.5_f)))),
// Now assign 'v' again with another frexp call.
// This requires generating a temporary variable for the struct initializer.
Assign("v", Call("frexp", vec3<f32>(4.5_a, 5.5_a, 6.5_a))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(struct frexp_result_vec3 {
float3 fract;
int3 exp;
};
[numthreads(1, 1, 1)]
void test_function() {
frexp_result_vec3 v = {float3(0.75f, 0.625f, 0.875f), int3(1, 2, 2)};
const frexp_result_vec3 c = {float3(0.5625f, 0.6875f, 0.8125f), (3).xxx};
v = c;
return;
}
)");
}
// TODO(crbug.com/tint/1757): Remove once deprecation period for `frexp().sig` is over
TEST_F(HlslGeneratorImplTest_Builtin, Frexp_Sig_Deprecation) {
WrapInFunction(Var("v", Call("frexp", 1_f)), //
MemberAccessor("v", "sig"));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(struct frexp_result {
float fract;
int exp;
};
[numthreads(1, 1, 1)]
void test_function() {
frexp_result v = {0.5f, 1};
const float tint_symbol = v.fract;
return;
}
)");

141
src/tint/writer/msl/generator_impl_builtin_test.cc
View File

@ -621,9 +621,9 @@ kernel void test_function() {
)");
}
TEST_F(MslGeneratorImplTest, Frexp_Scalar_f32) {
auto* call = Call("frexp", 1_f);
WrapInFunction(CallStmt(call));
TEST_F(MslGeneratorImplTest, Runtime_Frexp_Scalar_f32) {
WrapInFunction(Var("f", Expr(1_f)), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -643,18 +643,19 @@ frexp_result tint_frexp(float param_0) {
}
kernel void test_function() {
tint_frexp(1.0f);
float f = 1.0f;
frexp_result v = tint_frexp(f);
return;
}
)");
}
TEST_F(MslGeneratorImplTest, Frexp_Scalar_f16) {
TEST_F(MslGeneratorImplTest, Runtime_Frexp_Scalar_f16) {
Enable(ast::Extension::kF16);
auto* call = Call("frexp", 1_h);
WrapInFunction(CallStmt(call));
WrapInFunction(Var("f", Expr(1_h)), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -674,16 +675,17 @@ frexp_result_f16 tint_frexp(half param_0) {
}
kernel void test_function() {
tint_frexp(1.0h);
half f = 1.0h;
frexp_result_f16 v = tint_frexp(f);
return;
}
)");
}
TEST_F(MslGeneratorImplTest, Frexp_Vector_f32) {
auto* call = Call("frexp", vec3<f32>());
WrapInFunction(CallStmt(call));
TEST_F(MslGeneratorImplTest, Runtime_Frexp_Vector_f32) {
WrapInFunction(Var("f", Expr(vec3<f32>())), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -703,18 +705,19 @@ frexp_result_vec3 tint_frexp(float3 param_0) {
}
kernel void test_function() {
tint_frexp(float3(0.0f));
float3 f = float3(0.0f);
frexp_result_vec3 v = tint_frexp(f);
return;
}
)");
}
TEST_F(MslGeneratorImplTest, Frexp_Vector_f16) {
TEST_F(MslGeneratorImplTest, Runtime_Frexp_Vector_f16) {
Enable(ast::Extension::kF16);
auto* call = Call("frexp", vec3<f16>());
WrapInFunction(CallStmt(call));
WrapInFunction(Var("f", Expr(vec3<f16>())), //
Var("v", Call("frexp", "f")));
GeneratorImpl& gen = SanitizeAndBuild();
@ -734,7 +737,100 @@ frexp_result_vec3_f16 tint_frexp(half3 param_0) {
}
kernel void test_function() {
tint_frexp(half3(0.0h));
half3 f = half3(0.0h);
frexp_result_vec3_f16 v = tint_frexp(f);
return;
}
)");
}
TEST_F(MslGeneratorImplTest, Const_Frexp_Scalar_f32) {
WrapInFunction(Decl(Let("v", Call("frexp", 1_f))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
using namespace metal;
struct frexp_result {
float fract;
int exp;
};
kernel void test_function() {
frexp_result const v = frexp_result{.fract=0.5f, .exp=1};
return;
}
)");
}
TEST_F(MslGeneratorImplTest, Const_Frexp_Scalar_f16) {
Enable(ast::Extension::kF16);
WrapInFunction(Decl(Let("v", Call("frexp", 1_h))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
using namespace metal;
struct frexp_result_f16 {
half fract;
int exp;
};
kernel void test_function() {
frexp_result_f16 const v = frexp_result_f16{.fract=0.5h, .exp=1};
return;
}
)");
}
TEST_F(MslGeneratorImplTest, Const_Frexp_Vector_f32) {
WrapInFunction(Decl(Let("v", Call("frexp", vec3<f32>()))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
using namespace metal;
struct frexp_result_vec3 {
float3 fract;
int3 exp;
};
kernel void test_function() {
frexp_result_vec3 const v = frexp_result_vec3{};
return;
}
)");
}
TEST_F(MslGeneratorImplTest, Const_Frexp_Vector_f16) {
Enable(ast::Extension::kF16);
WrapInFunction(Decl(Let("v", Call("frexp", vec3<f16>()))));
GeneratorImpl& gen = SanitizeAndBuild();
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
using namespace metal;
struct frexp_result_vec3_f16 {
half3 fract;
int3 exp;
};
kernel void test_function() {
frexp_result_vec3_f16 const v = frexp_result_vec3_f16{};
return;
}
@ -751,19 +847,8 @@ TEST_F(MslGeneratorImplTest, Frexp_Sig_Deprecation) {
EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
using namespace metal;
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
kernel void test_function() {
float const tint_symbol = tint_frexp(1.0f).fract;
float const tint_symbol = 0.5f;
return;
}

166
src/tint/writer/spirv/builder_builtin_test.cc
View File

@ -1834,7 +1834,7 @@ OpFunctionEnd
Validate(b);
}
TEST_F(BuiltinBuilderTest, Call_Frexp_f32) {
TEST_F(BuiltinBuilderTest, Runtime_Call_Frexp_f32) {
auto* vec = Var("vec", vec2<f32>(1_f, 2_f));
auto* expr = Call("frexp", vec);
Func("a_func", utils::Empty, ty.void_(),
@ -1888,7 +1888,7 @@ OpFunctionEnd
Validate(b);
}
TEST_F(BuiltinBuilderTest, Call_Frexp_f16) {
TEST_F(BuiltinBuilderTest, Runtime_Call_Frexp_f16) {
Enable(ast::Extension::kF16);
auto* vec = Var("vec", vec2<f16>(1_h, 2_h));
@ -1948,15 +1948,10 @@ OpFunctionEnd
Validate(b);
}
// TODO(crbug.com/tint/1757): Remove once deprecation period for `frexp().sig` is over
TEST_F(BuiltinBuilderTest, Frexp_Sig_Deprecation) {
WrapInFunction(MemberAccessor(Call("frexp", 1_f), "sig"));
auto* vec = Var("vec", vec2<f32>(1_f, 2_f));
TEST_F(BuiltinBuilderTest, Const_Call_Frexp_f32) {
Func("a_func", utils::Empty, ty.void_(),
utils::Vector{
Decl(vec),
Decl(Let("s", MemberAccessor(Call("frexp", vec), "sig"))),
CallStmt(Call("frexp", vec2<f32>(1_f, 2_f))),
},
utils::Vector{
Stage(ast::PipelineStage::kFragment),
@ -1967,51 +1962,134 @@ TEST_F(BuiltinBuilderTest, Frexp_Sig_Deprecation) {
ASSERT_TRUE(b.Build()) << b.error();
auto got = DumpBuilder(b);
auto* expect = R"(OpCapability Shader
%9 = OpExtInstImport "GLSL.std.450"
%11 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %3 "test_function"
OpEntryPoint Fragment %12 "a_func"
OpExecutionMode %3 LocalSize 1 1 1
OpExecutionMode %12 OriginUpperLeft
OpName %3 "test_function"
OpName %6 "__frexp_result"
OpEntryPoint Fragment %3 "a_func"
OpExecutionMode %3 OriginUpperLeft
OpName %3 "a_func"
OpName %6 "__frexp_result_vec2"
OpMemberName %6 0 "fract"
OpMemberName %6 1 "exp"
OpName %12 "a_func"
OpName %17 "vec"
OpName %21 "__frexp_result_vec2"
OpMemberName %21 0 "fract"
OpMemberName %21 1 "exp"
OpMemberDecorate %6 0 Offset 0
OpMemberDecorate %6 1 Offset 4
OpMemberDecorate %21 0 Offset 0
OpMemberDecorate %21 1 Offset 8
OpMemberDecorate %6 1 Offset 8
%2 = OpTypeVoid
%1 = OpTypeFunction %2
%7 = OpTypeFloat 32
%8 = OpTypeInt 32 1
%6 = OpTypeStruct %7 %8
%10 = OpConstant %7 1
%14 = OpTypeVector %7 2
%15 = OpConstant %7 2
%16 = OpConstantComposite %14 %10 %15
%18 = OpTypePointer Function %14
%19 = OpConstantNull %14
%22 = OpTypeVector %8 2
%21 = OpTypeStruct %14 %22
%8 = OpTypeFloat 32
%7 = OpTypeVector %8 2
%10 = OpTypeInt 32 1
%9 = OpTypeVector %10 2
%6 = OpTypeStruct %7 %9
%12 = OpConstant %8 1
%13 = OpConstant %8 2
%14 = OpConstantComposite %7 %12 %13
%3 = OpFunction %2 None %1
%4 = OpLabel
%5 = OpExtInst %6 %9 FrexpStruct %10
%11 = OpCompositeExtract %7 %5 0
%5 = OpExtInst %6 %11 FrexpStruct %14
OpReturn
OpFunctionEnd
%12 = OpFunction %2 None %1
%13 = OpLabel
%17 = OpVariable %18 Function %19
OpStore %17 %16
%23 = OpLoad %14 %17
%20 = OpExtInst %21 %9 FrexpStruct %23
%24 = OpCompositeExtract %14 %20 0
)";
EXPECT_EQ(expect, got);
Validate(b);
}
TEST_F(BuiltinBuilderTest, Const_Call_Frexp_f16) {
Enable(ast::Extension::kF16);
Func("a_func", utils::Empty, ty.void_(),
utils::Vector{
CallStmt(Call("frexp", vec2<f16>(1_h, 2_h))),
},
utils::Vector{
Stage(ast::PipelineStage::kFragment),
});
spirv::Builder& b = Build();
ASSERT_TRUE(b.Build()) << b.error();
auto got = DumpBuilder(b);
auto* expect = R"(OpCapability Shader
OpCapability Float16
OpCapability UniformAndStorageBuffer16BitAccess
OpCapability StorageBuffer16BitAccess
OpCapability StorageInputOutput16
%11 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Fragment %3 "a_func"
OpExecutionMode %3 OriginUpperLeft
OpName %3 "a_func"
OpName %6 "__frexp_result_vec2_f16"
OpMemberName %6 0 "fract"
OpMemberName %6 1 "exp"
OpMemberDecorate %6 0 Offset 0
OpMemberDecorate %6 1 Offset 8
%2 = OpTypeVoid
%1 = OpTypeFunction %2
%8 = OpTypeFloat 16
%7 = OpTypeVector %8 2
%10 = OpTypeInt 32 1
%9 = OpTypeVector %10 2
%6 = OpTypeStruct %7 %9
%12 = OpConstant %8 0x1p+0
%13 = OpConstant %8 0x1p+1
%14 = OpConstantComposite %7 %12 %13
%3 = OpFunction %2 None %1
%4 = OpLabel
%5 = OpExtInst %6 %11 FrexpStruct %14
OpReturn
OpFunctionEnd
)";
EXPECT_EQ(expect, got);
Validate(b);
}
// TODO(crbug.com/tint/1757): Remove once deprecation period for `frexp().sig` is over
TEST_F(BuiltinBuilderTest, Frexp_Sig_Deprecation) {
Func("a_func", utils::Empty, ty.void_(),
utils::Vector{
Decl(Var("vec", vec2<f32>(1_f, 2_f))),
Decl(Let("s", MemberAccessor(Call("frexp", "vec"), "sig"))),
},
utils::Vector{
Stage(ast::PipelineStage::kFragment),
});
spirv::Builder& b = Build();
ASSERT_TRUE(b.Build()) << b.error();
auto got = DumpBuilder(b);
auto* expect = R"(OpCapability Shader
%17 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Fragment %3 "a_func"
OpExecutionMode %3 OriginUpperLeft
OpName %3 "a_func"
OpName %10 "vec"
OpName %14 "__frexp_result_vec2"
OpMemberName %14 0 "fract"
OpMemberName %14 1 "exp"
OpMemberDecorate %14 0 Offset 0
OpMemberDecorate %14 1 Offset 8
%2 = OpTypeVoid
%1 = OpTypeFunction %2
%6 = OpTypeFloat 32
%5 = OpTypeVector %6 2
%7 = OpConstant %6 1
%8 = OpConstant %6 2
%9 = OpConstantComposite %5 %7 %8
%11 = OpTypePointer Function %5
%12 = OpConstantNull %5
%16 = OpTypeInt 32 1
%15 = OpTypeVector %16 2
%14 = OpTypeStruct %5 %15
%3 = OpFunction %2 None %1
%4 = OpLabel
%10 = OpVariable %11 Function %12
OpStore %10 %9
%18 = OpLoad %5 %10
%13 = OpExtInst %14 %17 FrexpStruct %18
%19 = OpCompositeExtract %5 %13 0
OpReturn
OpFunctionEnd
)";

9
test/tint/bug/tint/1757.wgsl.expected.dxc.hlsl
View File

@ -6,16 +6,9 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const frexp_result res = tint_frexp(1.230000019f);
const frexp_result res = {0.61500001f, 1};
const int exp = res.exp;
const float sig = res.fract;
return;

9
test/tint/bug/tint/1757.wgsl.expected.fxc.hlsl
View File

@ -6,16 +6,9 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const frexp_result res = tint_frexp(1.230000019f);
const frexp_result res = {0.61500001f, 1};
const int exp = res.exp;
const float sig = res.fract;
return;

8
test/tint/bug/tint/1757.wgsl.expected.glsl
View File

@ -9,15 +9,9 @@ struct frexp_result {
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void tint_symbol() {
frexp_result res = tint_frexp(1.230000019f);
frexp_result res = frexp_result(0.61500001f, 1);
int tint_symbol_1 = res.exp;
float sig = res.fract;
}

8
test/tint/bug/tint/1757.wgsl.expected.msl
View File

@ -10,14 +10,8 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
kernel void tint_symbol() {
frexp_result const res = tint_frexp(1.230000019f);
frexp_result const res = frexp_result{.fract=0.61500001f, .exp=1};
int const exp = res.exp;
float const sig = res.fract;
return;

10
test/tint/bug/tint/1757.wgsl.expected.spvasm
View File

@ -8,7 +8,6 @@ bug/tint/1757.wgsl:6:25 warning: use of deprecated language feature: 'sig' has b
; Bound: 13
; Schema: 0
OpCapability Shader
%9 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
@ -23,11 +22,12 @@ bug/tint/1757.wgsl:6:25 warning: use of deprecated language feature: 'sig' has b
%float = OpTypeFloat 32
%int = OpTypeInt 32 1
%__frexp_result = OpTypeStruct %float %int
%float_1_23000002 = OpConstant %float 1.23000002
%float_0_61500001 = OpConstant %float 0.61500001
%int_1 = OpConstant %int 1
%10 = OpConstantComposite %__frexp_result %float_0_61500001 %int_1
%main = OpFunction %void None %1
%4 = OpLabel
%5 = OpExtInst %__frexp_result %9 FrexpStruct %float_1_23000002
%11 = OpCompositeExtract %int %5 1
%12 = OpCompositeExtract %float %5 0
%11 = OpCompositeExtract %int %10 1
%12 = OpCompositeExtract %float %10 0
OpReturn
OpFunctionEnd

9
test/tint/builtins/frexp.wgsl.expected.dxc.hlsl
View File

@ -2,16 +2,9 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const frexp_result res = tint_frexp(1.230000019f);
const frexp_result res = {0.61500001f, 1};
const int exp = res.exp;
const float fract = res.fract;
return;

9
test/tint/builtins/frexp.wgsl.expected.fxc.hlsl
View File

@ -2,16 +2,9 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const frexp_result res = tint_frexp(1.230000019f);
const frexp_result res = {0.61500001f, 1};
const int exp = res.exp;
const float fract = res.fract;
return;

8
test/tint/builtins/frexp.wgsl.expected.glsl
View File

@ -5,15 +5,9 @@ struct frexp_result {
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void tint_symbol() {
frexp_result res = tint_frexp(1.230000019f);
frexp_result res = frexp_result(0.61500001f, 1);
int tint_symbol_1 = res.exp;
float tint_symbol_2 = res.fract;
}

8
test/tint/builtins/frexp.wgsl.expected.msl
View File

@ -6,14 +6,8 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
kernel void tint_symbol() {
frexp_result const res = tint_frexp(1.230000019f);
frexp_result const res = frexp_result{.fract=0.61500001f, .exp=1};
int const exp = res.exp;
float const fract = res.fract;
return;

10
test/tint/builtins/frexp.wgsl.expected.spvasm
View File

@ -4,7 +4,6 @@
; Bound: 13
; Schema: 0
OpCapability Shader
%9 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
@ -19,11 +18,12 @@
%float = OpTypeFloat 32
%int = OpTypeInt 32 1
%__frexp_result = OpTypeStruct %float %int
%float_1_23000002 = OpConstant %float 1.23000002
%float_0_61500001 = OpConstant %float 0.61500001
%int_1 = OpConstant %int 1
%10 = OpConstantComposite %__frexp_result %float_0_61500001 %int_1
%main = OpFunction %void None %1
%4 = OpLabel
%5 = OpExtInst %__frexp_result %9 FrexpStruct %float_1_23000002
%11 = OpCompositeExtract %int %5 1
%12 = OpCompositeExtract %float %5 0
%11 = OpCompositeExtract %int %10 1
%12 = OpCompositeExtract %float %10 0
OpReturn
OpFunctionEnd

7
test/tint/builtins/frexp/scalar/const.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = 1.25;
let res = frexp(in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

11
test/tint/builtins/frexp/scalar/const.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,11 @@
struct frexp_result {
float fract;
int exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result res = {0.625f, 1};
const float fract = res.fract;
const int exp = res.exp;
return;
}

11
test/tint/builtins/frexp/scalar/const.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,11 @@
struct frexp_result {
float fract;
int exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result res = {0.625f, 1};
const float fract = res.fract;
const int exp = res.exp;
return;
}

19
test/tint/builtins/frexp/scalar/const.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,19 @@
#version 310 es
struct frexp_result {
float fract;
int exp;
};
void tint_symbol() {
frexp_result res = frexp_result(0.625f, 1);
float tint_symbol_2 = res.fract;
int tint_symbol_3 = res.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

15
test/tint/builtins/frexp/scalar/const.wgsl.expected.msl Normal file
View File

@ -0,0 +1,15 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result {
float fract;
int exp;
};
kernel void tint_symbol() {
frexp_result const res = frexp_result{.fract=0.625f, .exp=1};
float const fract = res.fract;
int const exp = res.exp;
return;
}

29
test/tint/builtins/frexp/scalar/const.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,29 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 13
; Schema: 0
OpCapability Shader
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
OpName %__frexp_result "__frexp_result"
OpMemberName %__frexp_result 0 "fract"
OpMemberName %__frexp_result 1 "exp"
OpMemberDecorate %__frexp_result 0 Offset 0
OpMemberDecorate %__frexp_result 1 Offset 4
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%int = OpTypeInt 32 1
%__frexp_result = OpTypeStruct %float %int
%float_0_625 = OpConstant %float 0.625
%int_1 = OpConstant %int 1
%10 = OpConstantComposite %__frexp_result %float_0_625 %int_1
%main = OpFunction %void None %1
%4 = OpLabel
%11 = OpCompositeExtract %float %10 0
%12 = OpCompositeExtract %int %10 1
OpReturn
OpFunctionEnd

7
test/tint/builtins/frexp/scalar/const.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = 1.25;
let res = frexp(in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

7
test/tint/builtins/frexp/scalar/const_members.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = 1.25;
const res = frexp(in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

12
test/tint/builtins/frexp/scalar/const_members.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,12 @@
struct frexp_result {
float fract;
int exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result tint_symbol_1 = {0.625f, 1};
const float fract = tint_symbol_1.fract;
const frexp_result tint_symbol_2 = {0.625f, 1};
const int exp = tint_symbol_2.exp;
return;
}

12
test/tint/builtins/frexp/scalar/const_members.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,12 @@
struct frexp_result {
float fract;
int exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result tint_symbol_1 = {0.625f, 1};
const float fract = tint_symbol_1.fract;
const frexp_result tint_symbol_2 = {0.625f, 1};
const int exp = tint_symbol_2.exp;
return;
}

20
test/tint/builtins/frexp/scalar/const_members.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,20 @@
#version 310 es
struct frexp_result {
float fract;
int exp;
};
void tint_symbol() {
frexp_result tint_symbol_4 = frexp_result(0.625f, 1);
float tint_symbol_2 = tint_symbol_4.fract;
frexp_result tint_symbol_5 = frexp_result(0.625f, 1);
int tint_symbol_3 = tint_symbol_5.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

16
test/tint/builtins/frexp/scalar/const_members.wgsl.expected.msl Normal file
View File

@ -0,0 +1,16 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result {
float fract;
int exp;
};
kernel void tint_symbol() {
frexp_result const tint_symbol_1 = frexp_result{.fract=0.625f, .exp=1};
float const fract = tint_symbol_1.fract;
frexp_result const tint_symbol_2 = frexp_result{.fract=0.625f, .exp=1};
int const exp = tint_symbol_2.exp;
return;
}

20
test/tint/builtins/frexp/scalar/const_members.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,20 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 9
; Schema: 0
OpCapability Shader
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%float_0_625 = OpConstant %float 0.625
%int = OpTypeInt 32 1
%int_1 = OpConstant %int 1
%main = OpFunction %void None %1
%4 = OpLabel
OpReturn
OpFunctionEnd

7
test/tint/builtins/frexp/scalar/const_members.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = 1.25;
const res = frexp(in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

10
test/tint/builtins/frexp/scalar/mixed.wgsl Normal file
View File

@ -0,0 +1,10 @@
@compute @workgroup_size(1)
fn main() {
const const_in = 1.25;
let runtime_in = 1.25;
var res = frexp(const_in);
res = frexp(runtime_in);
res = frexp(const_in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

22
test/tint/builtins/frexp/scalar/mixed.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,22 @@
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float runtime_in = 1.25f;
frexp_result res = {0.625f, 1};
res = tint_frexp(runtime_in);
const frexp_result c = {0.625f, 1};
res = c;
const float fract = res.fract;
const int exp = res.exp;
return;
}

22
test/tint/builtins/frexp/scalar/mixed.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,22 @@
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float runtime_in = 1.25f;
frexp_result res = {0.625f, 1};
res = tint_frexp(runtime_in);
const frexp_result c = {0.625f, 1};
res = c;
const float fract = res.fract;
const int exp = res.exp;
return;
}

28
test/tint/builtins/frexp/scalar/mixed.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,28 @@
#version 310 es
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void tint_symbol() {
float runtime_in = 1.25f;
frexp_result res = frexp_result(0.625f, 1);
res = tint_frexp(runtime_in);
res = frexp_result(0.625f, 1);
float tint_symbol_1 = res.fract;
int tint_symbol_2 = res.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

24
test/tint/builtins/frexp/scalar/mixed.wgsl.expected.msl Normal file
View File

@ -0,0 +1,24 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
kernel void tint_symbol() {
float const runtime_in = 1.25f;
frexp_result res = frexp_result{.fract=0.625f, .exp=1};
res = tint_frexp(runtime_in);
res = frexp_result{.fract=0.625f, .exp=1};
float const fract = res.fract;
int const exp = res.exp;
return;
}

46
test/tint/builtins/frexp/scalar/mixed.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,46 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 26
; Schema: 0
OpCapability Shader
%16 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
OpName %__frexp_result "__frexp_result"
OpMemberName %__frexp_result 0 "fract"
OpMemberName %__frexp_result 1 "exp"
OpName %res "res"
OpMemberDecorate %__frexp_result 0 Offset 0
OpMemberDecorate %__frexp_result 1 Offset 4
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%float_1_25 = OpConstant %float 1.25
%int = OpTypeInt 32 1
%__frexp_result = OpTypeStruct %float %int
%float_0_625 = OpConstant %float 0.625
%int_1 = OpConstant %int 1
%11 = OpConstantComposite %__frexp_result %float_0_625 %int_1
%_ptr_Function___frexp_result = OpTypePointer Function %__frexp_result
%14 = OpConstantNull %__frexp_result
%uint = OpTypeInt 32 0
%uint_0 = OpConstant %uint 0
%_ptr_Function_float = OpTypePointer Function %float
%uint_1 = OpConstant %uint 1
%_ptr_Function_int = OpTypePointer Function %int
%main = OpFunction %void None %1
%4 = OpLabel
%res = OpVariable %_ptr_Function___frexp_result Function %14
OpStore %res %11
%15 = OpExtInst %__frexp_result %16 FrexpStruct %float_1_25
OpStore %res %15
OpStore %res %11
%20 = OpAccessChain %_ptr_Function_float %res %uint_0
%21 = OpLoad %float %20
%24 = OpAccessChain %_ptr_Function_int %res %uint_1
%25 = OpLoad %int %24
OpReturn
OpFunctionEnd

10
test/tint/builtins/frexp/scalar/mixed.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,10 @@
@compute @workgroup_size(1)
fn main() {
const const_in = 1.25;
let runtime_in = 1.25;
var res = frexp(const_in);
res = frexp(runtime_in);
res = frexp(const_in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

7
test/tint/builtins/frexp/scalar/runtime.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
let in = 1.25;
let res = frexp(in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

19
test/tint/builtins/frexp/scalar/runtime.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,19 @@
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float tint_symbol = 1.25f;
const frexp_result res = tint_frexp(tint_symbol);
const float fract = res.fract;
const int exp = res.exp;
return;
}

19
test/tint/builtins/frexp/scalar/runtime.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,19 @@
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float tint_symbol = 1.25f;
const frexp_result res = tint_frexp(tint_symbol);
const float fract = res.fract;
const int exp = res.exp;
return;
}

26
test/tint/builtins/frexp/scalar/runtime.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,26 @@
#version 310 es
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void tint_symbol() {
float tint_symbol_1 = 1.25f;
frexp_result res = tint_frexp(tint_symbol_1);
float tint_symbol_2 = res.fract;
int tint_symbol_3 = res.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

22
test/tint/builtins/frexp/scalar/runtime.wgsl.expected.msl Normal file
View File

@ -0,0 +1,22 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
kernel void tint_symbol() {
float const in = 1.25f;
frexp_result const res = tint_frexp(in);
float const fract = res.fract;
int const exp = res.exp;
return;
}

29
test/tint/builtins/frexp/scalar/runtime.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,29 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 13
; Schema: 0
OpCapability Shader
%10 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
OpName %__frexp_result "__frexp_result"
OpMemberName %__frexp_result 0 "fract"
OpMemberName %__frexp_result 1 "exp"
OpMemberDecorate %__frexp_result 0 Offset 0
OpMemberDecorate %__frexp_result 1 Offset 4
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%float_1_25 = OpConstant %float 1.25
%int = OpTypeInt 32 1
%__frexp_result = OpTypeStruct %float %int
%main = OpFunction %void None %1
%4 = OpLabel
%7 = OpExtInst %__frexp_result %10 FrexpStruct %float_1_25
%11 = OpCompositeExtract %float %7 0
%12 = OpCompositeExtract %int %7 1
OpReturn
OpFunctionEnd

7
test/tint/builtins/frexp/scalar/runtime.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
let in = 1.25;
let res = frexp(in);
let fract : f32 = res.fract;
let exp : i32 = res.exp;
}

7
test/tint/builtins/frexp/vector/const.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = vec2(1.25, 3.75);
let res = frexp(in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

11
test/tint/builtins/frexp/vector/const.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,11 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result_vec2 res = {float2(0.625f, 0.9375f), int2(1, 2)};
const float2 fract = res.fract;
const int2 exp = res.exp;
return;
}

11
test/tint/builtins/frexp/vector/const.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,11 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result_vec2 res = {float2(0.625f, 0.9375f), int2(1, 2)};
const float2 fract = res.fract;
const int2 exp = res.exp;
return;
}

19
test/tint/builtins/frexp/vector/const.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,19 @@
#version 310 es
struct frexp_result_vec2 {
vec2 fract;
ivec2 exp;
};
void tint_symbol() {
frexp_result_vec2 res = frexp_result_vec2(vec2(0.625f, 0.9375f), ivec2(1, 2));
vec2 tint_symbol_2 = res.fract;
ivec2 tint_symbol_3 = res.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

15
test/tint/builtins/frexp/vector/const.wgsl.expected.msl Normal file
View File

@ -0,0 +1,15 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
kernel void tint_symbol() {
frexp_result_vec2 const res = frexp_result_vec2{.fract=float2(0.625f, 0.9375f), .exp=int2(1, 2)};
float2 const fract = res.fract;
int2 const exp = res.exp;
return;
}

35
test/tint/builtins/frexp/vector/const.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,35 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 19
; Schema: 0
OpCapability Shader
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
OpName %__frexp_result_vec2 "__frexp_result_vec2"
OpMemberName %__frexp_result_vec2 0 "fract"
OpMemberName %__frexp_result_vec2 1 "exp"
OpMemberDecorate %__frexp_result_vec2 0 Offset 0
OpMemberDecorate %__frexp_result_vec2 1 Offset 8
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%v2float = OpTypeVector %float 2
%int = OpTypeInt 32 1
%v2int = OpTypeVector %int 2
%__frexp_result_vec2 = OpTypeStruct %v2float %v2int
%float_0_625 = OpConstant %float 0.625
%float_0_9375 = OpConstant %float 0.9375
%12 = OpConstantComposite %v2float %float_0_625 %float_0_9375
%int_1 = OpConstant %int 1
%int_2 = OpConstant %int 2
%15 = OpConstantComposite %v2int %int_1 %int_2
%16 = OpConstantComposite %__frexp_result_vec2 %12 %15
%main = OpFunction %void None %1
%4 = OpLabel
%17 = OpCompositeExtract %v2float %16 0
%18 = OpCompositeExtract %v2int %16 1
OpReturn
OpFunctionEnd

7
test/tint/builtins/frexp/vector/const.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = vec2(1.25, 3.75);
let res = frexp(in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

7
test/tint/builtins/frexp/vector/const_members.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = vec2(1.25, 3.75);
const res = frexp(in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

12
test/tint/builtins/frexp/vector/const_members.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,12 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result_vec2 tint_symbol_1 = {float2(0.625f, 0.9375f), int2(1, 2)};
const float2 fract = tint_symbol_1.fract;
const frexp_result_vec2 tint_symbol_2 = {float2(0.625f, 0.9375f), int2(1, 2)};
const int2 exp = tint_symbol_2.exp;
return;
}

12
test/tint/builtins/frexp/vector/const_members.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,12 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
[numthreads(1, 1, 1)]
void main() {
const frexp_result_vec2 tint_symbol_1 = {float2(0.625f, 0.9375f), int2(1, 2)};
const float2 fract = tint_symbol_1.fract;
const frexp_result_vec2 tint_symbol_2 = {float2(0.625f, 0.9375f), int2(1, 2)};
const int2 exp = tint_symbol_2.exp;
return;
}

20
test/tint/builtins/frexp/vector/const_members.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,20 @@
#version 310 es
struct frexp_result_vec2 {
vec2 fract;
ivec2 exp;
};
void tint_symbol() {
frexp_result_vec2 tint_symbol_4 = frexp_result_vec2(vec2(0.625f, 0.9375f), ivec2(1, 2));
vec2 tint_symbol_2 = tint_symbol_4.fract;
frexp_result_vec2 tint_symbol_5 = frexp_result_vec2(vec2(0.625f, 0.9375f), ivec2(1, 2));
ivec2 tint_symbol_3 = tint_symbol_5.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

16
test/tint/builtins/frexp/vector/const_members.wgsl.expected.msl Normal file
View File

@ -0,0 +1,16 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
kernel void tint_symbol() {
frexp_result_vec2 const tint_symbol_1 = frexp_result_vec2{.fract=float2(0.625f, 0.9375f), .exp=int2(1, 2)};
float2 const fract = tint_symbol_1.fract;
frexp_result_vec2 const tint_symbol_2 = frexp_result_vec2{.fract=float2(0.625f, 0.9375f), .exp=int2(1, 2)};
int2 const exp = tint_symbol_2.exp;
return;
}

26
test/tint/builtins/frexp/vector/const_members.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,26 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 15
; Schema: 0
OpCapability Shader
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%v2float = OpTypeVector %float 2
%float_0_625 = OpConstant %float 0.625
%float_0_9375 = OpConstant %float 0.9375
%9 = OpConstantComposite %v2float %float_0_625 %float_0_9375
%int = OpTypeInt 32 1
%v2int = OpTypeVector %int 2
%int_1 = OpConstant %int 1
%int_2 = OpConstant %int 2
%14 = OpConstantComposite %v2int %int_1 %int_2
%main = OpFunction %void None %1
%4 = OpLabel
OpReturn
OpFunctionEnd

7
test/tint/builtins/frexp/vector/const_members.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
const in = vec2(1.25, 3.75);
const res = frexp(in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

10
test/tint/builtins/frexp/vector/mixed.wgsl Normal file
View File

@ -0,0 +1,10 @@
@compute @workgroup_size(1)
fn main() {
const const_in = vec2(1.25, 3.75);
let runtime_in = vec2(1.25, 3.75);
var res = frexp(const_in);
res = frexp(runtime_in);
res = frexp(const_in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

22
test/tint/builtins/frexp/vector/mixed.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,22 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
frexp_result_vec2 tint_frexp(float2 param_0) {
float2 exp;
float2 fract = frexp(param_0, exp);
frexp_result_vec2 result = {fract, int2(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float2 runtime_in = float2(1.25f, 3.75f);
frexp_result_vec2 res = {float2(0.625f, 0.9375f), int2(1, 2)};
res = tint_frexp(runtime_in);
const frexp_result_vec2 c = {float2(0.625f, 0.9375f), int2(1, 2)};
res = c;
const float2 fract = res.fract;
const int2 exp = res.exp;
return;
}

22
test/tint/builtins/frexp/vector/mixed.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,22 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
frexp_result_vec2 tint_frexp(float2 param_0) {
float2 exp;
float2 fract = frexp(param_0, exp);
frexp_result_vec2 result = {fract, int2(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float2 runtime_in = float2(1.25f, 3.75f);
frexp_result_vec2 res = {float2(0.625f, 0.9375f), int2(1, 2)};
res = tint_frexp(runtime_in);
const frexp_result_vec2 c = {float2(0.625f, 0.9375f), int2(1, 2)};
res = c;
const float2 fract = res.fract;
const int2 exp = res.exp;
return;
}

28
test/tint/builtins/frexp/vector/mixed.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,28 @@
#version 310 es
struct frexp_result_vec2 {
vec2 fract;
ivec2 exp;
};
frexp_result_vec2 tint_frexp(vec2 param_0) {
frexp_result_vec2 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void tint_symbol() {
vec2 runtime_in = vec2(1.25f, 3.75f);
frexp_result_vec2 res = frexp_result_vec2(vec2(0.625f, 0.9375f), ivec2(1, 2));
res = tint_frexp(runtime_in);
res = frexp_result_vec2(vec2(0.625f, 0.9375f), ivec2(1, 2));
vec2 tint_symbol_1 = res.fract;
ivec2 tint_symbol_2 = res.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

24
test/tint/builtins/frexp/vector/mixed.wgsl.expected.msl Normal file
View File

@ -0,0 +1,24 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
frexp_result_vec2 tint_frexp(float2 param_0) {
frexp_result_vec2 result;
result.fract = frexp(param_0, result.exp);
return result;
}
kernel void tint_symbol() {
float2 const runtime_in = float2(1.25f, 3.75f);
frexp_result_vec2 res = frexp_result_vec2{.fract=float2(0.625f, 0.9375f), .exp=int2(1, 2)};
res = tint_frexp(runtime_in);
res = frexp_result_vec2{.fract=float2(0.625f, 0.9375f), .exp=int2(1, 2)};
float2 const fract = res.fract;
int2 const exp = res.exp;
return;
}

54
test/tint/builtins/frexp/vector/mixed.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,54 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 34
; Schema: 0
OpCapability Shader
%24 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
OpName %__frexp_result_vec2 "__frexp_result_vec2"
OpMemberName %__frexp_result_vec2 0 "fract"
OpMemberName %__frexp_result_vec2 1 "exp"
OpName %res "res"
OpMemberDecorate %__frexp_result_vec2 0 Offset 0
OpMemberDecorate %__frexp_result_vec2 1 Offset 8
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%v2float = OpTypeVector %float 2
%float_1_25 = OpConstant %float 1.25
%float_3_75 = OpConstant %float 3.75
%9 = OpConstantComposite %v2float %float_1_25 %float_3_75
%int = OpTypeInt 32 1
%v2int = OpTypeVector %int 2
%__frexp_result_vec2 = OpTypeStruct %v2float %v2int
%float_0_625 = OpConstant %float 0.625
%float_0_9375 = OpConstant %float 0.9375
%15 = OpConstantComposite %v2float %float_0_625 %float_0_9375
%int_1 = OpConstant %int 1
%int_2 = OpConstant %int 2
%18 = OpConstantComposite %v2int %int_1 %int_2
%19 = OpConstantComposite %__frexp_result_vec2 %15 %18
%_ptr_Function___frexp_result_vec2 = OpTypePointer Function %__frexp_result_vec2
%22 = OpConstantNull %__frexp_result_vec2
%uint = OpTypeInt 32 0
%uint_0 = OpConstant %uint 0
%_ptr_Function_v2float = OpTypePointer Function %v2float
%uint_1 = OpConstant %uint 1
%_ptr_Function_v2int = OpTypePointer Function %v2int
%main = OpFunction %void None %1
%4 = OpLabel
%res = OpVariable %_ptr_Function___frexp_result_vec2 Function %22
OpStore %res %19
%23 = OpExtInst %__frexp_result_vec2 %24 FrexpStruct %9
OpStore %res %23
OpStore %res %19
%28 = OpAccessChain %_ptr_Function_v2float %res %uint_0
%29 = OpLoad %v2float %28
%32 = OpAccessChain %_ptr_Function_v2int %res %uint_1
%33 = OpLoad %v2int %32
OpReturn
OpFunctionEnd

10
test/tint/builtins/frexp/vector/mixed.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,10 @@
@compute @workgroup_size(1)
fn main() {
const const_in = vec2(1.25, 3.75);
let runtime_in = vec2(1.25, 3.75);
var res = frexp(const_in);
res = frexp(runtime_in);
res = frexp(const_in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

7
test/tint/builtins/frexp/vector/runtime.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
let in = vec2(1.25, 3.75);
let res = frexp(in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

19
test/tint/builtins/frexp/vector/runtime.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,19 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
frexp_result_vec2 tint_frexp(float2 param_0) {
float2 exp;
float2 fract = frexp(param_0, exp);
frexp_result_vec2 result = {fract, int2(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float2 tint_symbol = float2(1.25f, 3.75f);
const frexp_result_vec2 res = tint_frexp(tint_symbol);
const float2 fract = res.fract;
const int2 exp = res.exp;
return;
}

19
test/tint/builtins/frexp/vector/runtime.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,19 @@
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
frexp_result_vec2 tint_frexp(float2 param_0) {
float2 exp;
float2 fract = frexp(param_0, exp);
frexp_result_vec2 result = {fract, int2(exp)};
return result;
}
[numthreads(1, 1, 1)]
void main() {
const float2 tint_symbol = float2(1.25f, 3.75f);
const frexp_result_vec2 res = tint_frexp(tint_symbol);
const float2 fract = res.fract;
const int2 exp = res.exp;
return;
}

26
test/tint/builtins/frexp/vector/runtime.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,26 @@
#version 310 es
struct frexp_result_vec2 {
vec2 fract;
ivec2 exp;
};
frexp_result_vec2 tint_frexp(vec2 param_0) {
frexp_result_vec2 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void tint_symbol() {
vec2 tint_symbol_1 = vec2(1.25f, 3.75f);
frexp_result_vec2 res = tint_frexp(tint_symbol_1);
vec2 tint_symbol_2 = res.fract;
ivec2 tint_symbol_3 = res.exp;
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
tint_symbol();
return;
}

22
test/tint/builtins/frexp/vector/runtime.wgsl.expected.msl Normal file
View File

@ -0,0 +1,22 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result_vec2 {
float2 fract;
int2 exp;
};
frexp_result_vec2 tint_frexp(float2 param_0) {
frexp_result_vec2 result;
result.fract = frexp(param_0, result.exp);
return result;
}
kernel void tint_symbol() {
float2 const in = float2(1.25f, 3.75f);
frexp_result_vec2 const res = tint_frexp(in);
float2 const fract = res.fract;
int2 const exp = res.exp;
return;
}

33
test/tint/builtins/frexp/vector/runtime.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,33 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 17
; Schema: 0
OpCapability Shader
%14 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %main "main"
OpExecutionMode %main LocalSize 1 1 1
OpName %main "main"
OpName %__frexp_result_vec2 "__frexp_result_vec2"
OpMemberName %__frexp_result_vec2 0 "fract"
OpMemberName %__frexp_result_vec2 1 "exp"
OpMemberDecorate %__frexp_result_vec2 0 Offset 0
OpMemberDecorate %__frexp_result_vec2 1 Offset 8
%void = OpTypeVoid
%1 = OpTypeFunction %void
%float = OpTypeFloat 32
%v2float = OpTypeVector %float 2
%float_1_25 = OpConstant %float 1.25
%float_3_75 = OpConstant %float 3.75
%9 = OpConstantComposite %v2float %float_1_25 %float_3_75
%int = OpTypeInt 32 1
%v2int = OpTypeVector %int 2
%__frexp_result_vec2 = OpTypeStruct %v2float %v2int
%main = OpFunction %void None %1
%4 = OpLabel
%10 = OpExtInst %__frexp_result_vec2 %14 FrexpStruct %9
%15 = OpCompositeExtract %v2float %10 0
%16 = OpCompositeExtract %v2int %10 1
OpReturn
OpFunctionEnd

7
test/tint/builtins/frexp/vector/runtime.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,7 @@
@compute @workgroup_size(1)
fn main() {
let in = vec2(1.25, 3.75);
let res = frexp(in);
let fract : vec2<f32> = res.fract;
let exp : vec2<i32> = res.exp;
}

43
test/tint/builtins/gen/literal/frexp/34bbfb.wgsl Normal file
View File

@ -0,0 +1,43 @@
// Copyright 2022 The Tint Authors.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
////////////////////////////////////////////////////////////////////////////////
// File generated by tools/src/cmd/gen
// using the template:
// test/tint/builtins/gen/gen.wgsl.tmpl
//
// Do not modify this file directly
////////////////////////////////////////////////////////////////////////////////
// fn frexp(vec<4, fa>) -> __frexp_result_vec<4, fa>
fn frexp_34bbfb() {
var res = frexp(vec4(1.));
}
@vertex
fn vertex_main() -> @builtin(position) vec4<f32> {
frexp_34bbfb();
return vec4<f32>();
}
@fragment
fn fragment_main() {
frexp_34bbfb();
}
@compute @workgroup_size(1)
fn compute_main() {
frexp_34bbfb();
}

34
test/tint/builtins/gen/literal/frexp/34bbfb.wgsl.expected.dxc.hlsl Normal file
View File

@ -0,0 +1,34 @@
struct frexp_result_vec4 {
float4 fract;
int4 exp;
};
void frexp_34bbfb() {
frexp_result_vec4 res = {(0.5f).xxxx, (1).xxxx};
}
struct tint_symbol {
float4 value : SV_Position;
};
float4 vertex_main_inner() {
frexp_34bbfb();
return (0.0f).xxxx;
}
tint_symbol vertex_main() {
const float4 inner_result = vertex_main_inner();
tint_symbol wrapper_result = (tint_symbol)0;
wrapper_result.value = inner_result;
return wrapper_result;
}
void fragment_main() {
frexp_34bbfb();
return;
}
[numthreads(1, 1, 1)]
void compute_main() {
frexp_34bbfb();
return;
}

34
test/tint/builtins/gen/literal/frexp/34bbfb.wgsl.expected.fxc.hlsl Normal file
View File

@ -0,0 +1,34 @@
struct frexp_result_vec4 {
float4 fract;
int4 exp;
};
void frexp_34bbfb() {
frexp_result_vec4 res = {(0.5f).xxxx, (1).xxxx};
}
struct tint_symbol {
float4 value : SV_Position;
};
float4 vertex_main_inner() {
frexp_34bbfb();
return (0.0f).xxxx;
}
tint_symbol vertex_main() {
const float4 inner_result = vertex_main_inner();
tint_symbol wrapper_result = (tint_symbol)0;
wrapper_result.value = inner_result;
return wrapper_result;
}
void fragment_main() {
frexp_34bbfb();
return;
}
[numthreads(1, 1, 1)]
void compute_main() {
frexp_34bbfb();
return;
}

67
test/tint/builtins/gen/literal/frexp/34bbfb.wgsl.expected.glsl Normal file
View File

@ -0,0 +1,67 @@
#version 310 es
struct frexp_result_vec4 {
vec4 fract;
ivec4 exp;
};
void frexp_34bbfb() {
frexp_result_vec4 res = frexp_result_vec4(vec4(0.5f), ivec4(1));
}
vec4 vertex_main() {
frexp_34bbfb();
return vec4(0.0f);
}
void main() {
gl_PointSize = 1.0;
vec4 inner_result = vertex_main();
gl_Position = inner_result;
gl_Position.y = -(gl_Position.y);
gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w);
return;
}
#version 310 es
precision mediump float;
struct frexp_result_vec4 {
vec4 fract;
ivec4 exp;
};
void frexp_34bbfb() {
frexp_result_vec4 res = frexp_result_vec4(vec4(0.5f), ivec4(1));
}
void fragment_main() {
frexp_34bbfb();
}
void main() {
fragment_main();
return;
}
#version 310 es
struct frexp_result_vec4 {
vec4 fract;
ivec4 exp;
};
void frexp_34bbfb() {
frexp_result_vec4 res = frexp_result_vec4(vec4(0.5f), ivec4(1));
}
void compute_main() {
frexp_34bbfb();
}
layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
compute_main();
return;
}

38
test/tint/builtins/gen/literal/frexp/34bbfb.wgsl.expected.msl Normal file
View File

@ -0,0 +1,38 @@
#include <metal_stdlib>
using namespace metal;
struct frexp_result_vec4 {
float4 fract;
int4 exp;
};
void frexp_34bbfb() {
frexp_result_vec4 res = frexp_result_vec4{.fract=float4(0.5f), .exp=int4(1)};
}
struct tint_symbol {
float4 value [[position]];
};
float4 vertex_main_inner() {
frexp_34bbfb();
return float4(0.0f);
}
vertex tint_symbol vertex_main() {
float4 const inner_result = vertex_main_inner();
tint_symbol wrapper_result = {};
wrapper_result.value = inner_result;
return wrapper_result;
}
fragment void fragment_main() {
frexp_34bbfb();
return;
}
kernel void compute_main() {
frexp_34bbfb();
return;
}

77
test/tint/builtins/gen/literal/frexp/34bbfb.wgsl.expected.spvasm Normal file
View File

@ -0,0 +1,77 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 38
; Schema: 0
OpCapability Shader
OpMemoryModel Logical GLSL450
OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
OpEntryPoint Fragment %fragment_main "fragment_main"
OpEntryPoint GLCompute %compute_main "compute_main"
OpExecutionMode %fragment_main OriginUpperLeft
OpExecutionMode %compute_main LocalSize 1 1 1
OpName %value "value"
OpName %vertex_point_size "vertex_point_size"
OpName %frexp_34bbfb "frexp_34bbfb"
OpName %__frexp_result_vec4 "__frexp_result_vec4"
OpMemberName %__frexp_result_vec4 0 "fract"
OpMemberName %__frexp_result_vec4 1 "exp"
OpName %res "res"
OpName %vertex_main_inner "vertex_main_inner"
OpName %vertex_main "vertex_main"
OpName %fragment_main "fragment_main"
OpName %compute_main "compute_main"
OpDecorate %value BuiltIn Position
OpDecorate %vertex_point_size BuiltIn PointSize
OpMemberDecorate %__frexp_result_vec4 0 Offset 0
OpMemberDecorate %__frexp_result_vec4 1 Offset 16
%float = OpTypeFloat 32
%v4float = OpTypeVector %float 4
%_ptr_Output_v4float = OpTypePointer Output %v4float
%5 = OpConstantNull %v4float
%value = OpVariable %_ptr_Output_v4float Output %5
%_ptr_Output_float = OpTypePointer Output %float
%8 = OpConstantNull %float
%vertex_point_size = OpVariable %_ptr_Output_float Output %8
%void = OpTypeVoid
%9 = OpTypeFunction %void
%int = OpTypeInt 32 1
%v4int = OpTypeVector %int 4
%__frexp_result_vec4 = OpTypeStruct %v4float %v4int
%float_0_5 = OpConstant %float 0.5
%17 = OpConstantComposite %v4float %float_0_5 %float_0_5 %float_0_5 %float_0_5
%int_1 = OpConstant %int 1
%19 = OpConstantComposite %v4int %int_1 %int_1 %int_1 %int_1
%20 = OpConstantComposite %__frexp_result_vec4 %17 %19
%_ptr_Function___frexp_result_vec4 = OpTypePointer Function %__frexp_result_vec4
%23 = OpConstantNull %__frexp_result_vec4
%24 = OpTypeFunction %v4float
%float_1 = OpConstant %float 1
%frexp_34bbfb = OpFunction %void None %9
%12 = OpLabel
%res = OpVariable %_ptr_Function___frexp_result_vec4 Function %23
OpStore %res %20
OpReturn
OpFunctionEnd
%vertex_main_inner = OpFunction %v4float None %24
%26 = OpLabel
%27 = OpFunctionCall %void %frexp_34bbfb
OpReturnValue %5
OpFunctionEnd
%vertex_main = OpFunction %void None %9
%29 = OpLabel
%30 = OpFunctionCall %v4float %vertex_main_inner
OpStore %value %30
OpStore %vertex_point_size %float_1
OpReturn
OpFunctionEnd
%fragment_main = OpFunction %void None %9
%33 = OpLabel
%34 = OpFunctionCall %void %frexp_34bbfb
OpReturn
OpFunctionEnd
%compute_main = OpFunction %void None %9
%36 = OpLabel
%37 = OpFunctionCall %void %frexp_34bbfb
OpReturn
OpFunctionEnd

19
test/tint/builtins/gen/literal/frexp/34bbfb.wgsl.expected.wgsl Normal file
View File

@ -0,0 +1,19 @@
fn frexp_34bbfb() {
var res = frexp(vec4(1.0));
}
@vertex
fn vertex_main() -> @builtin(position) vec4<f32> {
frexp_34bbfb();
return vec4<f32>();
}
@fragment
fn fragment_main() {
frexp_34bbfb();
}
@compute @workgroup_size(1)
fn compute_main() {
frexp_34bbfb();
}

9
test/tint/builtins/gen/literal/frexp/3dd21e.wgsl.expected.dxc.hlsl
View File

@ -2,15 +2,8 @@ struct frexp_result_vec4_f16 {
vector<float16_t, 4> fract;
int4 exp;
};
frexp_result_vec4_f16 tint_frexp(vector<float16_t, 4> param_0) {
vector<float16_t, 4> exp;
vector<float16_t, 4> fract = frexp(param_0, exp);
frexp_result_vec4_f16 result = {fract, int4(exp)};
return result;
}
void frexp_3dd21e() {
frexp_result_vec4_f16 res = tint_frexp((float16_t(1.0h)).xxxx);
frexp_result_vec4_f16 res = {(float16_t(0.5h)).xxxx, (1).xxxx};
}
struct tint_symbol {

24
test/tint/builtins/gen/literal/frexp/3dd21e.wgsl.expected.glsl
View File

@ -6,15 +6,9 @@ struct frexp_result_vec4_f16 {
ivec4 exp;
};
frexp_result_vec4_f16 tint_frexp(f16vec4 param_0) {
frexp_result_vec4_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_3dd21e() {
frexp_result_vec4_f16 res = tint_frexp(f16vec4(1.0hf));
frexp_result_vec4_f16 res = frexp_result_vec4_f16(f16vec4(0.5hf), ivec4(1));
}
vec4 vertex_main() {
@ -39,15 +33,9 @@ struct frexp_result_vec4_f16 {
ivec4 exp;
};
frexp_result_vec4_f16 tint_frexp(f16vec4 param_0) {
frexp_result_vec4_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_3dd21e() {
frexp_result_vec4_f16 res = tint_frexp(f16vec4(1.0hf));
frexp_result_vec4_f16 res = frexp_result_vec4_f16(f16vec4(0.5hf), ivec4(1));
}
void fragment_main() {
@ -66,15 +54,9 @@ struct frexp_result_vec4_f16 {
ivec4 exp;
};
frexp_result_vec4_f16 tint_frexp(f16vec4 param_0) {
frexp_result_vec4_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_3dd21e() {
frexp_result_vec4_f16 res = tint_frexp(f16vec4(1.0hf));
frexp_result_vec4_f16 res = frexp_result_vec4_f16(f16vec4(0.5hf), ivec4(1));
}
void compute_main() {

8
test/tint/builtins/gen/literal/frexp/3dd21e.wgsl.expected.msl
View File

@ -6,14 +6,8 @@ struct frexp_result_vec4_f16 {
half4 fract;
int4 exp;
};
frexp_result_vec4_f16 tint_frexp(half4 param_0) {
frexp_result_vec4_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_3dd21e() {
frexp_result_vec4_f16 res = tint_frexp(half4(1.0h));
frexp_result_vec4_f16 res = frexp_result_vec4_f16{.fract=half4(0.5h), .exp=int4(1)};
}
struct tint_symbol {

39
test/tint/builtins/gen/literal/frexp/3dd21e.wgsl.expected.spvasm
View File

@ -1,14 +1,13 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 39
; Bound: 40
; Schema: 0
OpCapability Shader
OpCapability Float16
OpCapability UniformAndStorageBuffer16BitAccess
OpCapability StorageBuffer16BitAccess
OpCapability StorageInputOutput16
%19 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
OpEntryPoint Fragment %fragment_main "fragment_main"
@ -45,38 +44,40 @@
%int = OpTypeInt 32 1
%v4int = OpTypeVector %int 4
%__frexp_result_vec4_f16 = OpTypeStruct %v4half %v4int
%half_0x1p_0 = OpConstant %half 0x1p+0
%21 = OpConstantComposite %v4half %half_0x1p_0 %half_0x1p_0 %half_0x1p_0 %half_0x1p_0
%half_0x1pn1 = OpConstant %half 0x1p-1
%19 = OpConstantComposite %v4half %half_0x1pn1 %half_0x1pn1 %half_0x1pn1 %half_0x1pn1
%int_1 = OpConstant %int 1
%21 = OpConstantComposite %v4int %int_1 %int_1 %int_1 %int_1
%22 = OpConstantComposite %__frexp_result_vec4_f16 %19 %21
%_ptr_Function___frexp_result_vec4_f16 = OpTypePointer Function %__frexp_result_vec4_f16
%24 = OpConstantNull %__frexp_result_vec4_f16
%25 = OpTypeFunction %v4float
%25 = OpConstantNull %__frexp_result_vec4_f16
%26 = OpTypeFunction %v4float
%float_1 = OpConstant %float 1
%frexp_3dd21e = OpFunction %void None %9
%12 = OpLabel
%res = OpVariable %_ptr_Function___frexp_result_vec4_f16 Function %24
%13 = OpExtInst %__frexp_result_vec4_f16 %19 FrexpStruct %21
OpStore %res %13
%res = OpVariable %_ptr_Function___frexp_result_vec4_f16 Function %25
OpStore %res %22
OpReturn
OpFunctionEnd
%vertex_main_inner = OpFunction %v4float None %25
%27 = OpLabel
%28 = OpFunctionCall %void %frexp_3dd21e
%vertex_main_inner = OpFunction %v4float None %26
%28 = OpLabel
%29 = OpFunctionCall %void %frexp_3dd21e
OpReturnValue %5
OpFunctionEnd
%vertex_main = OpFunction %void None %9
%30 = OpLabel
%31 = OpFunctionCall %v4float %vertex_main_inner
OpStore %value %31
%31 = OpLabel
%32 = OpFunctionCall %v4float %vertex_main_inner
OpStore %value %32
OpStore %vertex_point_size %float_1
OpReturn
OpFunctionEnd
%fragment_main = OpFunction %void None %9
%34 = OpLabel
%35 = OpFunctionCall %void %frexp_3dd21e
%35 = OpLabel
%36 = OpFunctionCall %void %frexp_3dd21e
OpReturn
OpFunctionEnd
%compute_main = OpFunction %void None %9
%37 = OpLabel
%38 = OpFunctionCall %void %frexp_3dd21e
%38 = OpLabel
%39 = OpFunctionCall %void %frexp_3dd21e
OpReturn
OpFunctionEnd

9
test/tint/builtins/gen/literal/frexp/4b2200.wgsl.expected.dxc.hlsl
View File

@ -2,15 +2,8 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
void frexp_4b2200() {
frexp_result res = tint_frexp(1.0f);
frexp_result res = {0.5f, 1};
}
struct tint_symbol {

9
test/tint/builtins/gen/literal/frexp/4b2200.wgsl.expected.fxc.hlsl
View File

@ -2,15 +2,8 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
float exp;
float fract = frexp(param_0, exp);
frexp_result result = {fract, int(exp)};
return result;
}
void frexp_4b2200() {
frexp_result res = tint_frexp(1.0f);
frexp_result res = {0.5f, 1};
}
struct tint_symbol {

24
test/tint/builtins/gen/literal/frexp/4b2200.wgsl.expected.glsl
View File

@ -5,15 +5,9 @@ struct frexp_result {
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_4b2200() {
frexp_result res = tint_frexp(1.0f);
frexp_result res = frexp_result(0.5f, 1);
}
vec4 vertex_main() {
@ -37,15 +31,9 @@ struct frexp_result {
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_4b2200() {
frexp_result res = tint_frexp(1.0f);
frexp_result res = frexp_result(0.5f, 1);
}
void fragment_main() {
@ -63,15 +51,9 @@ struct frexp_result {
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_4b2200() {
frexp_result res = tint_frexp(1.0f);
frexp_result res = frexp_result(0.5f, 1);
}
void compute_main() {

8
test/tint/builtins/gen/literal/frexp/4b2200.wgsl.expected.msl
View File

@ -6,14 +6,8 @@ struct frexp_result {
float fract;
int exp;
};
frexp_result tint_frexp(float param_0) {
frexp_result result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_4b2200() {
frexp_result res = tint_frexp(1.0f);
frexp_result res = frexp_result{.fract=0.5f, .exp=1};
}
struct tint_symbol {

19
test/tint/builtins/gen/literal/frexp/4b2200.wgsl.expected.spvasm
View File

@ -1,10 +1,9 @@
; SPIR-V
; Version: 1.3
; Generator: Google Tint Compiler; 0
; Bound: 34
; Bound: 35
; Schema: 0
OpCapability Shader
%16 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
OpEntryPoint Fragment %fragment_main "fragment_main"
@ -38,15 +37,17 @@
%9 = OpTypeFunction %void
%int = OpTypeInt 32 1
%__frexp_result = OpTypeStruct %float %int
%float_1 = OpConstant %float 1
%float_0_5 = OpConstant %float 0.5
%int_1 = OpConstant %int 1
%17 = OpConstantComposite %__frexp_result %float_0_5 %int_1
%_ptr_Function___frexp_result = OpTypePointer Function %__frexp_result
%20 = OpConstantNull %__frexp_result
%21 = OpTypeFunction %v4float
%float_1 = OpConstant %float 1
%frexp_4b2200 = OpFunction %void None %9
%12 = OpLabel
%res = OpVariable %_ptr_Function___frexp_result Function %20
%13 = OpExtInst %__frexp_result %16 FrexpStruct %float_1
OpStore %res %13
OpStore %res %17
OpReturn
OpFunctionEnd
%vertex_main_inner = OpFunction %v4float None %21
@ -62,12 +63,12 @@
OpReturn
OpFunctionEnd
%fragment_main = OpFunction %void None %9
%29 = OpLabel
%30 = OpFunctionCall %void %frexp_4b2200
%30 = OpLabel
%31 = OpFunctionCall %void %frexp_4b2200
OpReturn
OpFunctionEnd
%compute_main = OpFunction %void None %9
%32 = OpLabel
%33 = OpFunctionCall %void %frexp_4b2200
%33 = OpLabel
%34 = OpFunctionCall %void %frexp_4b2200
OpReturn
OpFunctionEnd

9
test/tint/builtins/gen/literal/frexp/5257dd.wgsl.expected.dxc.hlsl
View File

@ -2,15 +2,8 @@ struct frexp_result_f16 {
float16_t fract;
int exp;
};
frexp_result_f16 tint_frexp(float16_t param_0) {
float16_t exp;
float16_t fract = frexp(param_0, exp);
frexp_result_f16 result = {fract, int(exp)};
return result;
}
void frexp_5257dd() {
frexp_result_f16 res = tint_frexp(float16_t(1.0h));
frexp_result_f16 res = {float16_t(0.5h), 1};
}
struct tint_symbol {

24
test/tint/builtins/gen/literal/frexp/5257dd.wgsl.expected.glsl
View File

@ -6,15 +6,9 @@ struct frexp_result_f16 {
int exp;
};
frexp_result_f16 tint_frexp(float16_t param_0) {
frexp_result_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_5257dd() {
frexp_result_f16 res = tint_frexp(1.0hf);
frexp_result_f16 res = frexp_result_f16(0.5hf, 1);
}
vec4 vertex_main() {
@ -39,15 +33,9 @@ struct frexp_result_f16 {
int exp;
};
frexp_result_f16 tint_frexp(float16_t param_0) {
frexp_result_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_5257dd() {
frexp_result_f16 res = tint_frexp(1.0hf);
frexp_result_f16 res = frexp_result_f16(0.5hf, 1);
}
void fragment_main() {
@ -66,15 +54,9 @@ struct frexp_result_f16 {
int exp;
};
frexp_result_f16 tint_frexp(float16_t param_0) {
frexp_result_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_5257dd() {
frexp_result_f16 res = tint_frexp(1.0hf);
frexp_result_f16 res = frexp_result_f16(0.5hf, 1);
}
void compute_main() {

8
test/tint/builtins/gen/literal/frexp/5257dd.wgsl.expected.msl
View File

@ -6,14 +6,8 @@ struct frexp_result_f16 {
half fract;
int exp;
};
frexp_result_f16 tint_frexp(half param_0) {
frexp_result_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_5257dd() {
frexp_result_f16 res = tint_frexp(1.0h);
frexp_result_f16 res = frexp_result_f16{.fract=0.5h, .exp=1};
}
struct tint_symbol {

8
test/tint/builtins/gen/literal/frexp/5257dd.wgsl.expected.spvasm
View File

@ -8,7 +8,6 @@
OpCapability UniformAndStorageBuffer16BitAccess
OpCapability StorageBuffer16BitAccess
OpCapability StorageInputOutput16
%17 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size
OpEntryPoint Fragment %fragment_main "fragment_main"
@ -43,7 +42,9 @@
%half = OpTypeFloat 16
%int = OpTypeInt 32 1
%__frexp_result_f16 = OpTypeStruct %half %int
%half_0x1p_0 = OpConstant %half 0x1p+0
%half_0x1pn1 = OpConstant %half 0x1p-1
%int_1 = OpConstant %int 1
%18 = OpConstantComposite %__frexp_result_f16 %half_0x1pn1 %int_1
%_ptr_Function___frexp_result_f16 = OpTypePointer Function %__frexp_result_f16
%21 = OpConstantNull %__frexp_result_f16
%22 = OpTypeFunction %v4float
@ -51,8 +52,7 @@
%frexp_5257dd = OpFunction %void None %9
%12 = OpLabel
%res = OpVariable %_ptr_Function___frexp_result_f16 Function %21
%13 = OpExtInst %__frexp_result_f16 %17 FrexpStruct %half_0x1p_0
OpStore %res %13
OpStore %res %18
OpReturn
OpFunctionEnd
%vertex_main_inner = OpFunction %v4float None %22

9
test/tint/builtins/gen/literal/frexp/5f47bf.wgsl.expected.dxc.hlsl
View File

@ -2,15 +2,8 @@ struct frexp_result_vec2_f16 {
vector<float16_t, 2> fract;
int2 exp;
};
frexp_result_vec2_f16 tint_frexp(vector<float16_t, 2> param_0) {
vector<float16_t, 2> exp;
vector<float16_t, 2> fract = frexp(param_0, exp);
frexp_result_vec2_f16 result = {fract, int2(exp)};
return result;
}
void frexp_5f47bf() {
frexp_result_vec2_f16 res = tint_frexp((float16_t(1.0h)).xx);
frexp_result_vec2_f16 res = {(float16_t(0.5h)).xx, (1).xx};
}
struct tint_symbol {

24
test/tint/builtins/gen/literal/frexp/5f47bf.wgsl.expected.glsl
View File

@ -6,15 +6,9 @@ struct frexp_result_vec2_f16 {
ivec2 exp;
};
frexp_result_vec2_f16 tint_frexp(f16vec2 param_0) {
frexp_result_vec2_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_5f47bf() {
frexp_result_vec2_f16 res = tint_frexp(f16vec2(1.0hf));
frexp_result_vec2_f16 res = frexp_result_vec2_f16(f16vec2(0.5hf), ivec2(1));
}
vec4 vertex_main() {
@ -39,15 +33,9 @@ struct frexp_result_vec2_f16 {
ivec2 exp;
};
frexp_result_vec2_f16 tint_frexp(f16vec2 param_0) {
frexp_result_vec2_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_5f47bf() {
frexp_result_vec2_f16 res = tint_frexp(f16vec2(1.0hf));
frexp_result_vec2_f16 res = frexp_result_vec2_f16(f16vec2(0.5hf), ivec2(1));
}
void fragment_main() {
@ -66,15 +54,9 @@ struct frexp_result_vec2_f16 {
ivec2 exp;
};
frexp_result_vec2_f16 tint_frexp(f16vec2 param_0) {
frexp_result_vec2_f16 result;
result.fract = frexp(param_0, result.exp);
return result;
}
void frexp_5f47bf() {
frexp_result_vec2_f16 res = tint_frexp(f16vec2(1.0hf));
frexp_result_vec2_f16 res = frexp_result_vec2_f16(f16vec2(0.5hf), ivec2(1));
}
void compute_main() {

Some files were not shown because too many files have changed in this diff Show More