Add ast::InternalDecoration
An tint-internal decoration used to add metadata between a sanitizer transform and a backend. Will be used for declaring backend-specific intrinsic calls. Change-Id: Ia05ba7dada0148de2d490605ba4d15c593075356 Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/46868 Commit-Queue: Ben Clayton <bclayton@google.com> Reviewed-by: James Price <jrprice@google.com>
This commit is contained in:
parent
86c2cbfb7e
commit
b502fdf82b
|
@ -281,6 +281,8 @@ source_set("libtint_core_src") {
|
||||||
"ast/if_statement.h",
|
"ast/if_statement.h",
|
||||||
"ast/int_literal.cc",
|
"ast/int_literal.cc",
|
||||||
"ast/int_literal.h",
|
"ast/int_literal.h",
|
||||||
|
"ast/internal_decoration.cc",
|
||||||
|
"ast/internal_decoration.h",
|
||||||
"ast/literal.cc",
|
"ast/literal.cc",
|
||||||
"ast/literal.h",
|
"ast/literal.h",
|
||||||
"ast/location_decoration.cc",
|
"ast/location_decoration.cc",
|
||||||
|
|
|
@ -96,6 +96,8 @@ set(TINT_LIB_SRCS
|
||||||
ast/if_statement.h
|
ast/if_statement.h
|
||||||
ast/int_literal.cc
|
ast/int_literal.cc
|
||||||
ast/int_literal.h
|
ast/int_literal.h
|
||||||
|
ast/internal_decoration.cc
|
||||||
|
ast/internal_decoration.h
|
||||||
ast/literal.cc
|
ast/literal.cc
|
||||||
ast/literal.h
|
ast/literal.h
|
||||||
ast/location_decoration.cc
|
ast/location_decoration.cc
|
||||||
|
|
|
@ -40,7 +40,6 @@ Function::Function(const Source& source,
|
||||||
for (auto* param : params_) {
|
for (auto* param : params_) {
|
||||||
TINT_ASSERT(param);
|
TINT_ASSERT(param);
|
||||||
}
|
}
|
||||||
TINT_ASSERT(body_);
|
|
||||||
TINT_ASSERT(symbol_.IsValid());
|
TINT_ASSERT(symbol_.IsValid());
|
||||||
TINT_ASSERT(return_type_);
|
TINT_ASSERT(return_type_);
|
||||||
}
|
}
|
||||||
|
|
|
@ -63,6 +63,18 @@ class Function : public Castable<Function, Node> {
|
||||||
/// @returns the decorations attached to this function
|
/// @returns the decorations attached to this function
|
||||||
const DecorationList& decorations() const { return decorations_; }
|
const DecorationList& decorations() const { return decorations_; }
|
||||||
|
|
||||||
|
/// @returns the decoration with the type `T` or nullptr if this function does
|
||||||
|
/// not contain a decoration with the given type
|
||||||
|
template <typename T>
|
||||||
|
const T* find_decoration() const {
|
||||||
|
for (auto* deco : decorations()) {
|
||||||
|
if (auto* d = deco->As<T>()) {
|
||||||
|
return d;
|
||||||
|
}
|
||||||
|
}
|
||||||
|
return nullptr;
|
||||||
|
}
|
||||||
|
|
||||||
/// @returns the workgroup size {x, y, z} for the function. {1, 1, 1} will be
|
/// @returns the workgroup size {x, y, z} for the function. {1, 1, 1} will be
|
||||||
/// return if no workgroup size was set.
|
/// return if no workgroup size was set.
|
||||||
std::tuple<uint32_t, uint32_t, uint32_t> workgroup_size() const;
|
std::tuple<uint32_t, uint32_t, uint32_t> workgroup_size() const;
|
||||||
|
|
|
@ -0,0 +1,34 @@
|
||||||
|
// Copyright 2021 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.
|
||||||
|
|
||||||
|
#include "src/ast/internal_decoration.h"
|
||||||
|
|
||||||
|
TINT_INSTANTIATE_TYPEINFO(tint::ast::InternalDecoration);
|
||||||
|
|
||||||
|
namespace tint {
|
||||||
|
namespace ast {
|
||||||
|
|
||||||
|
InternalDecoration::InternalDecoration() : Base(Source{}) {}
|
||||||
|
|
||||||
|
InternalDecoration::~InternalDecoration() = default;
|
||||||
|
|
||||||
|
void InternalDecoration::to_str(const semantic::Info&,
|
||||||
|
std::ostream& out,
|
||||||
|
size_t indent) const {
|
||||||
|
make_indent(out, indent);
|
||||||
|
out << "tint_internal(" << Name() << ")" << std::endl;
|
||||||
|
}
|
||||||
|
|
||||||
|
} // namespace ast
|
||||||
|
} // namespace tint
|
|
@ -0,0 +1,53 @@
|
||||||
|
// Copyright 2021 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.
|
||||||
|
|
||||||
|
#ifndef SRC_AST_INTERNAL_DECORATION_H_
|
||||||
|
#define SRC_AST_INTERNAL_DECORATION_H_
|
||||||
|
|
||||||
|
#include <string>
|
||||||
|
|
||||||
|
#include "src/ast/decoration.h"
|
||||||
|
|
||||||
|
namespace tint {
|
||||||
|
namespace ast {
|
||||||
|
|
||||||
|
/// A decoration used to indicate that a function is tint-internal.
|
||||||
|
/// These decorations are not produced by generators, but instead are usually
|
||||||
|
/// created by transforms for consumption by a particular backend.
|
||||||
|
/// Functions annotated with this decoration will have relaxed validation.
|
||||||
|
class InternalDecoration : public Castable<InternalDecoration, Decoration> {
|
||||||
|
public:
|
||||||
|
/// Constructor
|
||||||
|
InternalDecoration();
|
||||||
|
|
||||||
|
/// Destructor
|
||||||
|
~InternalDecoration() override;
|
||||||
|
|
||||||
|
/// @return a short description of the internal decoration which will be
|
||||||
|
/// displayed in WGSL as `[[internal(<name>)]]` (but is not parsable).
|
||||||
|
virtual std::string Name() const = 0;
|
||||||
|
|
||||||
|
/// Writes a representation of the node to the output stream
|
||||||
|
/// @param sem the semantic info for the program
|
||||||
|
/// @param out the stream to write to
|
||||||
|
/// @param indent number of spaces to indent the node when writing
|
||||||
|
void to_str(const semantic::Info& sem,
|
||||||
|
std::ostream& out,
|
||||||
|
size_t indent) const override;
|
||||||
|
};
|
||||||
|
|
||||||
|
} // namespace ast
|
||||||
|
} // namespace tint
|
||||||
|
|
||||||
|
#endif // SRC_AST_INTERNAL_DECORATION_H_
|
|
@ -26,6 +26,7 @@
|
||||||
#include "src/ast/discard_statement.h"
|
#include "src/ast/discard_statement.h"
|
||||||
#include "src/ast/fallthrough_statement.h"
|
#include "src/ast/fallthrough_statement.h"
|
||||||
#include "src/ast/if_statement.h"
|
#include "src/ast/if_statement.h"
|
||||||
|
#include "src/ast/internal_decoration.h"
|
||||||
#include "src/ast/loop_statement.h"
|
#include "src/ast/loop_statement.h"
|
||||||
#include "src/ast/return_statement.h"
|
#include "src/ast/return_statement.h"
|
||||||
#include "src/ast/struct_block_decoration.h"
|
#include "src/ast/struct_block_decoration.h"
|
||||||
|
@ -301,12 +302,18 @@ bool Resolver::ValidateFunction(const ast::Function* func) {
|
||||||
}
|
}
|
||||||
|
|
||||||
if (!func->return_type()->Is<type::Void>()) {
|
if (!func->return_type()->Is<type::Void>()) {
|
||||||
if (!func->get_last_statement() ||
|
if (func->body()) {
|
||||||
!func->get_last_statement()->Is<ast::ReturnStatement>()) {
|
if (!func->get_last_statement() ||
|
||||||
diagnostics_.add_error(
|
!func->get_last_statement()->Is<ast::ReturnStatement>()) {
|
||||||
"v-0002", "non-void function must end with a return statement",
|
diagnostics_.add_error(
|
||||||
func->source());
|
"v-0002", "non-void function must end with a return statement",
|
||||||
return false;
|
func->source());
|
||||||
|
return false;
|
||||||
|
}
|
||||||
|
} else if (!func->find_decoration<ast::InternalDecoration>()) {
|
||||||
|
TINT_ICE(diagnostics_)
|
||||||
|
<< "Function " << builder_->Symbols().NameFor(func->symbol())
|
||||||
|
<< " has no body and does not have the [[internal]] decoration";
|
||||||
}
|
}
|
||||||
|
|
||||||
for (auto* deco : func->return_type_decorations()) {
|
for (auto* deco : func->return_type_decorations()) {
|
||||||
|
@ -594,8 +601,10 @@ bool Resolver::Function(ast::Function* func) {
|
||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
if (!BlockStatement(func->body())) {
|
if (func->body()) {
|
||||||
return false;
|
if (!BlockStatement(func->body())) {
|
||||||
|
return false;
|
||||||
|
}
|
||||||
}
|
}
|
||||||
variable_stack_.pop_scope();
|
variable_stack_.pop_scope();
|
||||||
|
|
||||||
|
|
|
@ -20,6 +20,7 @@
|
||||||
#include "src/ast/call_statement.h"
|
#include "src/ast/call_statement.h"
|
||||||
#include "src/ast/constant_id_decoration.h"
|
#include "src/ast/constant_id_decoration.h"
|
||||||
#include "src/ast/fallthrough_statement.h"
|
#include "src/ast/fallthrough_statement.h"
|
||||||
|
#include "src/ast/internal_decoration.h"
|
||||||
#include "src/ast/variable_decl_statement.h"
|
#include "src/ast/variable_decl_statement.h"
|
||||||
#include "src/semantic/array.h"
|
#include "src/semantic/array.h"
|
||||||
#include "src/semantic/call.h"
|
#include "src/semantic/call.h"
|
||||||
|
@ -1464,6 +1465,11 @@ bool GeneratorImpl::EmitFunction(std::ostream& out, ast::Function* func) {
|
||||||
|
|
||||||
auto* func_sem = builder_.Sem().Get(func);
|
auto* func_sem = builder_.Sem().Get(func);
|
||||||
|
|
||||||
|
if (func->find_decoration<ast::InternalDecoration>()) {
|
||||||
|
// An internal function. Do not emit.
|
||||||
|
return true;
|
||||||
|
}
|
||||||
|
|
||||||
// TODO(dsinclair): This could be smarter. If the input/outputs for multiple
|
// TODO(dsinclair): This could be smarter. If the input/outputs for multiple
|
||||||
// entry points are the same we could generate a single struct and then have
|
// entry points are the same we could generate a single struct and then have
|
||||||
// this determine it's the same struct and just emit once.
|
// this determine it's the same struct and just emit once.
|
||||||
|
|
|
@ -21,6 +21,7 @@
|
||||||
#include "src/ast/call_statement.h"
|
#include "src/ast/call_statement.h"
|
||||||
#include "src/ast/constant_id_decoration.h"
|
#include "src/ast/constant_id_decoration.h"
|
||||||
#include "src/ast/float_literal.h"
|
#include "src/ast/float_literal.h"
|
||||||
|
#include "src/ast/internal_decoration.h"
|
||||||
#include "src/ast/module.h"
|
#include "src/ast/module.h"
|
||||||
#include "src/ast/sint_literal.h"
|
#include "src/ast/sint_literal.h"
|
||||||
#include "src/ast/stage_decoration.h"
|
#include "src/ast/stage_decoration.h"
|
||||||
|
@ -299,6 +300,9 @@ bool GeneratorImpl::EmitFunction(ast::Function* func) {
|
||||||
if (auto* stage = deco->As<ast::StageDecoration>()) {
|
if (auto* stage = deco->As<ast::StageDecoration>()) {
|
||||||
out_ << "stage(" << stage->value() << ")";
|
out_ << "stage(" << stage->value() << ")";
|
||||||
}
|
}
|
||||||
|
if (auto* internal = deco->As<ast::InternalDecoration>()) {
|
||||||
|
out_ << "internal(" << internal->Name() << ")";
|
||||||
|
}
|
||||||
out_ << "]]" << std::endl;
|
out_ << "]]" << std::endl;
|
||||||
}
|
}
|
||||||
|
|
||||||
|
@ -339,8 +343,14 @@ bool GeneratorImpl::EmitFunction(ast::Function* func) {
|
||||||
return false;
|
return false;
|
||||||
}
|
}
|
||||||
|
|
||||||
out_ << " ";
|
if (func->body()) {
|
||||||
return EmitBlockAndNewline(func->body());
|
out_ << " ";
|
||||||
|
return EmitBlockAndNewline(func->body());
|
||||||
|
} else {
|
||||||
|
out_ << std::endl;
|
||||||
|
}
|
||||||
|
|
||||||
|
return true;
|
||||||
}
|
}
|
||||||
|
|
||||||
bool GeneratorImpl::EmitImageFormat(const type::ImageFormat fmt) {
|
bool GeneratorImpl::EmitImageFormat(const type::ImageFormat fmt) {
|
||||||
|
|
Loading…
Reference in New Issue