diff --git a/build.zig b/build.zig index d69dcfad..2807cdab 100644 --- a/build.zig +++ b/build.zig @@ -9,6 +9,7 @@ const sysjs = @import("libs/sysjs/build.zig"); const earcut = @import("libs/earcut/build.zig"); const gamemode = @import("libs/gamemode/build.zig"); const model3d = @import("libs/model3d/build.zig"); +const dusk = @import("libs/dusk/build.zig"); const wasmserve = @import("tools/wasmserve/wasmserve.zig"); const gpu_dawn = @import("libs/gpu-dawn/sdk.zig").Sdk(.{ .glfw_include_dir = sdkPath("/libs/glfw/upstream/glfw/include"), @@ -66,6 +67,7 @@ pub fn build(b: *std.Build) !void { const basisu_test_step = b.step("test-basisu", "Run Basis-Universal library tests"); const sysaudio_test_step = b.step("test-sysaudio", "Run sysaudio library tests"); const model3d_test_step = b.step("test-model3d", "Run Model3D library tests"); + const dusk_test_step = b.step("test-dusk", "Run Dusk library tests"); const mach_test_step = b.step("test-mach", "Run Engine library tests"); core_test_step.dependOn(&(try core.testStep(b, optimize, target)).step); @@ -74,6 +76,7 @@ pub fn build(b: *std.Build) !void { basisu_test_step.dependOn(&basisu.testStep(b, optimize, target).step); sysaudio_test_step.dependOn(&sysaudio.testStep(b, optimize, target).step); model3d_test_step.dependOn(&model3d.testStep(b, optimize, target).step); + dusk_test_step.dependOn(&dusk.testStep(b, optimize, target).step); mach_test_step.dependOn(&testStep(b, optimize, target).step); all_tests_step.dependOn(core_test_step); @@ -82,6 +85,7 @@ pub fn build(b: *std.Build) !void { all_tests_step.dependOn(freetype_test_step); all_tests_step.dependOn(sysaudio_test_step); all_tests_step.dependOn(model3d_test_step); + all_tests_step.dependOn(dusk_test_step); all_tests_step.dependOn(mach_test_step); const shaderexp_app = try App.init( diff --git a/libs/dusk/.gitattributes b/libs/dusk/.gitattributes new file mode 100644 index 00000000..ba1273bd --- /dev/null +++ b/libs/dusk/.gitattributes @@ -0,0 +1,2 @@ +* text=auto eol=lf +upstream/** linguist-vendored diff --git a/libs/dusk/.github/FUNDING.yml b/libs/dusk/.github/FUNDING.yml new file mode 100644 index 00000000..99fd1066 --- /dev/null +++ b/libs/dusk/.github/FUNDING.yml @@ -0,0 +1 @@ +github: slimsag diff --git a/libs/dusk/.github/pull_request_template.md b/libs/dusk/.github/pull_request_template.md new file mode 100644 index 00000000..6230c782 --- /dev/null +++ b/libs/dusk/.github/pull_request_template.md @@ -0,0 +1,5 @@ +Please send your change to [the main repository](https://github.com/hexops/mach/tree/main/libs/dusk) instead, sorry for the trouble! + +This helps us avoid some complex merge conflicts we run into when changes are made to both repositories and history needs to be reconciled. Keeping PRs in just that repository enables us to use `git subtree` to trivially keep the two repositories in sync. + +Once your PR is merged over there, it'll automatically sync to this repository. diff --git a/libs/dusk/.github/workflows/ci.yml b/libs/dusk/.github/workflows/ci.yml new file mode 100644 index 00000000..8dfc0f18 --- /dev/null +++ b/libs/dusk/.github/workflows/ci.yml @@ -0,0 +1,16 @@ +name: CI +on: + - push + - pull_request +jobs: + x86_64-linux: + runs-on: ubuntu-latest + steps: + - name: Checkout + uses: actions/checkout@v2 + - name: Setup Zig + run: | + sudo apt install xz-utils + sudo sh -c 'wget -c https://ziglang.org/builds/zig-linux-x86_64-0.11.0-dev.1605+abc9530a8.tar.xz -O - | tar -xJ --strip-components=1 -C /usr/local/bin' + - name: test + run: zig build test diff --git a/libs/dusk/.gitignore b/libs/dusk/.gitignore new file mode 100644 index 00000000..feda423c --- /dev/null +++ b/libs/dusk/.gitignore @@ -0,0 +1,18 @@ +# This file is for zig-specific build artifacts. +# If you have OS-specific or editor-specific files to ignore, +# such as *.swp or .DS_Store, put those in your global +# ~/.gitignore and put this in your ~/.gitconfig: +# +# [core] +# excludesfile = ~/.gitignore +# +# Cheers! +# -andrewrk + +zig-cache/ +zig-out/ +/release/ +/debug/ +/build/ +/build-*/ +/docgen_tmp/ diff --git a/libs/dusk/LICENSE b/libs/dusk/LICENSE new file mode 100644 index 00000000..999b2584 --- /dev/null +++ b/libs/dusk/LICENSE @@ -0,0 +1,13 @@ +Copyright 2023, Hexops Contributors (given via the Git commit history). + +All documentation, image, sound, font, and 2D/3D model files are CC-BY-4.0 licensed unless +otherwise noted. You may get a copy of this license at https://creativecommons.org/licenses/by/4.0 + +Files in a directory with a separate LICENSE file may contain files under different license terms, +described within that LICENSE file. + +All other files are licensed under the Apache License, Version 2.0 (see LICENSE-APACHE or http://www.apache.org/licenses/LICENSE-2.0) +or the MIT license (see LICENSE-MIT or http://opensource.org/licenses/MIT), at your option. + +All files in the project without exclusions may not be copied, modified, or distributed except +according to the terms above. \ No newline at end of file diff --git a/libs/dusk/LICENSE-APACHE b/libs/dusk/LICENSE-APACHE new file mode 100644 index 00000000..d6456956 --- /dev/null +++ b/libs/dusk/LICENSE-APACHE @@ -0,0 +1,202 @@ + + Apache License + Version 2.0, January 2004 + http://www.apache.org/licenses/ + + TERMS AND CONDITIONS FOR USE, REPRODUCTION, AND DISTRIBUTION + + 1. Definitions. + + "License" shall mean the terms and conditions for use, reproduction, + and distribution as defined by Sections 1 through 9 of this document. + + "Licensor" shall mean the copyright owner or entity authorized by + the copyright owner that is granting the License. + + "Legal Entity" shall mean the union of the acting entity and all + other entities that control, are controlled by, or are under common + control with that entity. For the purposes of this definition, + "control" means (i) the power, direct or indirect, to cause the + direction or management of such entity, whether by contract or + otherwise, or (ii) ownership of fifty percent (50%) or more of the + outstanding shares, or (iii) beneficial ownership of such entity. + + "You" (or "Your") shall mean an individual or Legal Entity + exercising permissions granted by this License. + + "Source" form shall mean the preferred form for making modifications, + including but not limited to software source code, documentation + source, and configuration files. + + "Object" form shall mean any form resulting from mechanical + transformation or translation of a Source form, including but + not limited to compiled object code, generated documentation, + and conversions to other media types. + + "Work" shall mean the work of authorship, whether in Source or + Object form, made available under the License, as indicated by a + copyright notice that is included in or attached to the work + (an example is provided in the Appendix below). + + "Derivative Works" shall mean any work, whether in Source or Object + form, that is based on (or derived from) the Work and for which the + editorial revisions, annotations, elaborations, or other modifications + represent, as a whole, an original work of authorship. For the purposes + of this License, Derivative Works shall not include works that remain + separable from, or merely link (or bind by name) to the interfaces of, + the Work and Derivative Works thereof. + + "Contribution" shall mean any work of authorship, including + the original version of the Work and any modifications or additions + to that Work or Derivative Works thereof, that is intentionally + submitted to Licensor for inclusion in the Work by the copyright owner + or by an individual or Legal Entity authorized to submit on behalf of + the copyright owner. For the purposes of this definition, "submitted" + means any form of electronic, verbal, or written communication sent + to the Licensor or its representatives, including but not limited to + communication on electronic mailing lists, source code control systems, + and issue tracking systems that are managed by, or on behalf of, the + Licensor for the purpose of discussing and improving the Work, but + excluding communication that is conspicuously marked or otherwise + designated in writing by the copyright owner as "Not a Contribution." + + "Contributor" shall mean Licensor and any individual or Legal Entity + on behalf of whom a Contribution has been received by Licensor and + subsequently incorporated within the Work. + + 2. Grant of Copyright License. Subject to the terms and conditions of + this License, each Contributor hereby grants to You a perpetual, + worldwide, non-exclusive, no-charge, royalty-free, irrevocable + copyright license to reproduce, prepare Derivative Works of, + publicly display, publicly perform, sublicense, and distribute the + Work and such Derivative Works in Source or Object form. + + 3. Grant of Patent License. Subject to the terms and conditions of + this License, each Contributor hereby grants to You a perpetual, + worldwide, non-exclusive, no-charge, royalty-free, irrevocable + (except as stated in this section) patent license to make, have made, + use, offer to sell, sell, import, and otherwise transfer the Work, + where such license applies only to those patent claims licensable + by such Contributor that are necessarily infringed by their + Contribution(s) alone or by combination of their Contribution(s) + with the Work to which such Contribution(s) was submitted. If You + institute patent litigation against any entity (including a + cross-claim or counterclaim in a lawsuit) alleging that the Work + or a Contribution incorporated within the Work constitutes direct + or contributory patent infringement, then any patent licenses + granted to You under this License for that Work shall terminate + as of the date such litigation is filed. + + 4. Redistribution. You may reproduce and distribute copies of the + Work or Derivative Works thereof in any medium, with or without + modifications, and in Source or Object form, provided that You + meet the following conditions: + + (a) You must give any other recipients of the Work or + Derivative Works a copy of this License; and + + (b) You must cause any modified files to carry prominent notices + stating that You changed the files; and + + (c) You must retain, in the Source form of any Derivative Works + that You distribute, all copyright, patent, trademark, and + attribution notices from the Source form of the Work, + excluding those notices that do not pertain to any part of + the Derivative Works; and + + (d) If the Work includes a "NOTICE" text file as part of its + distribution, then any Derivative Works that You distribute must + include a readable copy of the attribution notices contained + within such NOTICE file, excluding those notices that do not + pertain to any part of the Derivative Works, in at least one + of the following places: within a NOTICE text file distributed + as part of the Derivative Works; within the Source form or + documentation, if provided along with the Derivative Works; or, + within a display generated by the Derivative Works, if and + wherever such third-party notices normally appear. The contents + of the NOTICE file are for informational purposes only and + do not modify the License. You may add Your own attribution + notices within Derivative Works that You distribute, alongside + or as an addendum to the NOTICE text from the Work, provided + that such additional attribution notices cannot be construed + as modifying the License. + + You may add Your own copyright statement to Your modifications and + may provide additional or different license terms and conditions + for use, reproduction, or distribution of Your modifications, or + for any such Derivative Works as a whole, provided Your use, + reproduction, and distribution of the Work otherwise complies with + the conditions stated in this License. + + 5. Submission of Contributions. Unless You explicitly state otherwise, + any Contribution intentionally submitted for inclusion in the Work + by You to the Licensor shall be under the terms and conditions of + this License, without any additional terms or conditions. + Notwithstanding the above, nothing herein shall supersede or modify + the terms of any separate license agreement you may have executed + with Licensor regarding such Contributions. + + 6. Trademarks. This License does not grant permission to use the trade + names, trademarks, service marks, or product names of the Licensor, + except as required for reasonable and customary use in describing the + origin of the Work and reproducing the content of the NOTICE file. + + 7. Disclaimer of Warranty. Unless required by applicable law or + agreed to in writing, Licensor provides the Work (and each + Contributor provides its Contributions) on an "AS IS" BASIS, + WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or + implied, including, without limitation, any warranties or conditions + of TITLE, NON-INFRINGEMENT, MERCHANTABILITY, or FITNESS FOR A + PARTICULAR PURPOSE. You are solely responsible for determining the + appropriateness of using or redistributing the Work and assume any + risks associated with Your exercise of permissions under this License. + + 8. Limitation of Liability. In no event and under no legal theory, + whether in tort (including negligence), contract, or otherwise, + unless required by applicable law (such as deliberate and grossly + negligent acts) or agreed to in writing, shall any Contributor be + liable to You for damages, including any direct, indirect, special, + incidental, or consequential damages of any character arising as a + result of this License or out of the use or inability to use the + Work (including but not limited to damages for loss of goodwill, + work stoppage, computer failure or malfunction, or any and all + other commercial damages or losses), even if such Contributor + has been advised of the possibility of such damages. + + 9. Accepting Warranty or Additional Liability. While redistributing + the Work or Derivative Works thereof, You may choose to offer, + and charge a fee for, acceptance of support, warranty, indemnity, + or other liability obligations and/or rights consistent with this + License. However, in accepting such obligations, You may act only + on Your own behalf and on Your sole responsibility, not on behalf + of any other Contributor, and only if You agree to indemnify, + defend, and hold each Contributor harmless for any liability + incurred by, or claims asserted against, such Contributor by reason + of your accepting any such warranty or additional liability. + + END OF TERMS AND CONDITIONS + + APPENDIX: How to apply the Apache License to your work. + + To apply the Apache License to your work, attach the following + boilerplate notice, with the fields enclosed by brackets "[]" + replaced with your own identifying information. (Don't include + the brackets!) The text should be enclosed in the appropriate + comment syntax for the file format. We also recommend that a + file or class name and description of purpose be included on the + same "printed page" as the copyright notice for easier + identification within third-party archives. + + Copyright [yyyy] [name of copyright owner] + + 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. diff --git a/libs/dusk/LICENSE-MIT b/libs/dusk/LICENSE-MIT new file mode 100644 index 00000000..14cd7973 --- /dev/null +++ b/libs/dusk/LICENSE-MIT @@ -0,0 +1,25 @@ +Copyright (c) 2021 Hexops Contributors (given via the Git commit history). + +Permission is hereby granted, free of charge, to any +person obtaining a copy of this software and associated +documentation files (the "Software"), to deal in the +Software without restriction, including without +limitation the rights to use, copy, modify, merge, +publish, distribute, sublicense, and/or sell copies of +the Software, and to permit persons to whom the Software +is furnished to do so, subject to the following +conditions: + +The above copyright notice and this permission notice +shall be included in all copies or substantial portions +of the Software. + +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF +ANY KIND, EXPRESS OR IMPLIED, INCLUDING BUT NOT LIMITED +TO THE WARRANTIES OF MERCHANTABILITY, FITNESS FOR A +PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT +SHALL THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY +CLAIM, DAMAGES OR OTHER LIABILITY, WHETHER IN AN ACTION +OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR +IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER +DEALINGS IN THE SOFTWARE. diff --git a/libs/dusk/README.md b/libs/dusk/README.md new file mode 100644 index 00000000..6ea2b00a --- /dev/null +++ b/libs/dusk/README.md @@ -0,0 +1,29 @@ +# mach/dusk - WebGPU implementation in Zig + +This repository is a separate copy of the same library in the [main Mach repository](https://github.com/hexops/mach), and is automatically kept in sync, so that anyone can use this library in their own project if they like! + +## Current Status + +Dusk is in **very early stages** and under heavy development; there are hundreds of known bugs/missing features. + +### WGSL compiler + +- [x] Parser +- [ ] Ast analysis (WIP!) +- [ ] Transpilation targets + - [ ] GLSL + - [ ] Spir-V + - [ ] HLSL + - [ ] Metal + +## Join the community + +Join the Mach community [on Discord](https://discord.gg/XNG3NZgCqp) to discuss this project, ask questions, get help, etc. + +## Issues + +Issues are tracked in the [main Mach repository](https://github.com/hexops/mach/issues?q=is%3Aissue+is%3Aopen+label%3Adusk). + +## Contributing + +Contributions are very welcome. Pull requests must be sent to [the main repository](https://github.com/hexops/mach/tree/main/libs/dusk) to avoid some complex merge conflicts we'd get by accepting contributions in both repositories. Once the changes are merged there, they'll get sync'd to this repository automatically. diff --git a/libs/dusk/build.zig b/libs/dusk/build.zig new file mode 100644 index 00000000..e195758a --- /dev/null +++ b/libs/dusk/build.zig @@ -0,0 +1,34 @@ +const std = @import("std"); + +pub fn build(b: *std.Build) !void { + const optimize = b.standardOptimizeOption(.{}); + const target = b.standardTargetOptions(.{}); + + const test_step = b.step("test", "Run library tests"); + test_step.dependOn(&testStep(b, optimize, target).step); +} + +pub fn module(b: *std.build.Builder) *std.build.Module { + return b.createModule(.{ .source_file = .{ .path = sdkPath("/src/main.zig") } }); +} + +pub fn testStep(b: *std.Build, optimize: std.builtin.OptimizeMode, target: std.zig.CrossTarget) *std.build.RunStep { + const main_tests = b.addTest(.{ + .name = "dusk-tests", + .kind = .test_exe, + .root_source_file = .{ .path = sdkPath("/test/main.zig") }, + .target = target, + .optimize = optimize, + }); + main_tests.addModule("dusk", module(b)); + main_tests.install(); + return main_tests.run(); +} + +fn sdkPath(comptime suffix: []const u8) []const u8 { + if (suffix[0] != '/') @compileError("suffix must be an absolute path"); + return comptime blk: { + const root_dir = std.fs.path.dirname(@src().file) orelse "."; + break :blk root_dir ++ suffix; + }; +} diff --git a/libs/dusk/src/Analyse.zig b/libs/dusk/src/Analyse.zig new file mode 100644 index 00000000..8dd29d32 --- /dev/null +++ b/libs/dusk/src/Analyse.zig @@ -0,0 +1,143 @@ +const std = @import("std"); +const Ast = @import("Ast.zig"); +const Token = @import("Token.zig"); +const ErrorMsg = @import("main.zig").ErrorMsg; +const Analyse = @This(); + +allocator: std.mem.Allocator, +tree: *const Ast, +errors: std.ArrayListUnmanaged(ErrorMsg), + +pub fn deinit(self: *Analyse) void { + for (self.errors.items) |*err_msg| err_msg.deinit(self.allocator); + self.errors.deinit(self.allocator); +} + +pub fn analyseRoot(self: *Analyse) !void { + const global_items = self.tree.spanToList(0); + + for (global_items, 0..) |node_i, i| { + try self.checkRedeclaration(global_items[i + 1 ..], node_i); + try self.globalDecl(global_items, node_i); + } + + if (self.errors.items.len > 0) { + return error.Analysing; + } +} + +pub fn globalDecl(self: *Analyse, parent_scope: []const Ast.Index, node_i: Ast.Index) !void { + switch (self.tree.nodeTag(node_i)) { + .global_variable => {}, // TODO + .struct_decl => try self.structDecl(parent_scope, node_i), + else => std.debug.print("Global Decl TODO: {}\n", .{self.tree.nodeTag(node_i)}), + } +} + +pub fn structDecl(self: *Analyse, parent_scope: []const Ast.Index, node: Ast.Index) !void { + const member_list = self.tree.spanToList(self.tree.nodeLHS(node)); + for (member_list, 0..) |member_node, i| { + try self.checkRedeclaration(member_list[i + 1 ..], member_node); + + const member_loc = self.tree.tokenLoc(self.tree.nodeToken(member_node)); + const member_name = member_loc.slice(self.tree.source); + const member_type_node = self.tree.nodeRHS(member_node); + + switch (self.tree.nodeTag(member_type_node)) { + .scalar_type, + .vector_type, + .matrix_type, + .atomic_type, + => {}, + .array_type => { + if (self.tree.nodeRHS(member_type_node) == Ast.null_index and + i != member_list.len - 1) + { + try self.addError( + member_loc, + "struct member with runtime-sized array type, must be the last member of the structure", + .{}, + null, + ); + } + }, + .user_type => { + _ = self.findDeclNode(parent_scope, member_name) orelse { + try self.addError( + member_loc, // TODO + "use of undeclared identifier '{s}'", + .{member_name}, + null, + ); + continue; + }; + }, + else => { + try self.addError( + member_loc, + "invalid struct member type '{s}'", + .{member_name}, + null, + ); + }, + } + } +} + +pub fn findDeclNode(self: *Analyse, scope_items: []const Ast.Index, name: []const u8) ?Ast.Index { + for (scope_items) |node| { + const node_token = self.declNameToken(node) orelse continue; + if (std.mem.eql(u8, name, self.tree.tokenLoc(node_token).slice(self.tree.source))) { + return node; + } + } + return null; +} + +pub fn checkRedeclaration(self: *Analyse, scope_items: []const Ast.Index, decl_node: Ast.Index) !void { + const decl_token_loc = self.tree.tokenLoc(self.declNameToken(decl_node).?); + const decl_name = decl_token_loc.slice(self.tree.source); + for (scope_items) |redecl_node| { + std.debug.assert(decl_node != redecl_node); + const redecl_token_loc = self.tree.tokenLoc(self.declNameToken(redecl_node).?); + const redecl_name = redecl_token_loc.slice(self.tree.source); + if (std.mem.eql(u8, decl_name, redecl_name)) { + try self.addError( + redecl_token_loc, + "redeclaration of '{s}'", + .{decl_name}, + try ErrorMsg.Note.create( + self.allocator, + decl_token_loc, + "other declaration here", + .{}, + ), + ); + } + } +} + +pub fn declNameToken(self: *Analyse, node: Ast.Index) ?Ast.Index { + return switch (self.tree.nodeTag(node)) { + .global_variable => self.tree.extraData(Ast.Node.GlobalVarDecl, self.tree.nodeLHS(node)).name, + .struct_decl, + .fn_decl, + .global_constant, + .override, + .type_alias, + => self.tree.nodeToken(node) + 1, + .struct_member => self.tree.nodeToken(node), + else => null, + }; +} + +pub fn addError( + self: *Analyse, + loc: Token.Loc, + comptime format: []const u8, + args: anytype, + note: ?ErrorMsg.Note, +) !void { + const err_msg = try ErrorMsg.create(self.allocator, loc, format, args, note); + try self.errors.append(self.allocator, err_msg); +} diff --git a/libs/dusk/src/Ast.zig b/libs/dusk/src/Ast.zig new file mode 100644 index 00000000..f8b82c1b --- /dev/null +++ b/libs/dusk/src/Ast.zig @@ -0,0 +1,715 @@ +const std = @import("std"); +const Analyse = @import("Analyse.zig"); +const Parser = @import("Parser.zig"); +const Token = @import("Token.zig"); +const Tokenizer = @import("Tokenizer.zig"); +const ErrorMsg = @import("main.zig").ErrorMsg; +const Extension = @import("main.zig").Extension; + +const Ast = @This(); + +pub const NodeList = std.MultiArrayList(Node); +pub const TokenList = std.MultiArrayList(Token); + +source: [:0]const u8, +tokens: TokenList.Slice, +nodes: NodeList.Slice, +extra: []const Index, + +pub fn deinit(tree: *Ast, allocator: std.mem.Allocator) void { + tree.tokens.deinit(allocator); + tree.nodes.deinit(allocator); + allocator.free(tree.extra); + tree.* = undefined; +} + +pub const ParseResult = union(enum) { + errors: []ErrorMsg, + tree: Ast, +}; + +/// parses a TranslationUnit (WGSL Program) +pub fn parse(allocator: std.mem.Allocator, source: [:0]const u8) !ParseResult { + const estimated_tokens = source.len / 8; + var tokens = std.MultiArrayList(Token){}; + try tokens.ensureTotalCapacity(allocator, estimated_tokens); + + var tokenizer = Tokenizer.init(source); + while (true) { + const tok = tokenizer.next(); + try tokens.append(allocator, tok); + if (tok.tag == .eof) break; + } + + var p = Parser{ + .allocator = allocator, + .source = source, + .tok_i = 0, + .tokens = tokens.toOwnedSlice(), + .nodes = .{}, + .extra = .{}, + .scratch = .{}, + .errors = .{}, + .extensions = Extension.Array.initFill(false), + }; + defer p.deinit(); + errdefer p.tokens.deinit(allocator); + + // TODO: make sure tokens:nodes ratio is right + const estimated_node_count = (tokens.len + 2) / 2; + try p.nodes.ensureTotalCapacity(allocator, estimated_node_count); + + p.parseRoot() catch |err| { + if (err == error.Parsing) { + p.tokens.deinit(allocator); + return .{ .errors = try p.errors.toOwnedSlice(allocator) }; + } + return err; + }; + + return .{ + .tree = .{ + .source = source, + .tokens = p.tokens, + .nodes = p.nodes.toOwnedSlice(), + .extra = try p.extra.toOwnedSlice(allocator), + }, + }; +} + +pub fn analyse(tree: Ast, allocator: std.mem.Allocator) !?[]ErrorMsg { + var analyser = Analyse{ + .allocator = allocator, + .tree = &tree, + .errors = .{}, + }; + defer analyser.deinit(); + + analyser.analyseRoot() catch |err| { + if (err == error.Analysing) { + return try analyser.errors.toOwnedSlice(allocator); + } + return err; + }; + + return null; +} + +pub fn spanToList(tree: Ast, span: Ast.Index) []const Ast.Index { + std.debug.assert(tree.nodeTag(span) == .span); + return tree.extra[tree.nodeLHS(span)..tree.nodeRHS(span)]; +} + +pub fn extraData(tree: Ast, comptime T: type, index: Ast.Index) T { + const fields = std.meta.fields(T); + var result: T = undefined; + inline for (fields, 0..) |field, i| { + comptime std.debug.assert(field.type == Ast.Index); + @field(result, field.name) = tree.extra[index + i]; + } + return result; +} + +pub fn tokenTag(tree: Ast, i: Index) Token.Tag { + return tree.tokens.items(.tag)[i]; +} + +pub fn tokenLoc(tree: Ast, i: Index) Token.Loc { + return tree.tokens.items(.loc)[i]; +} + +pub fn nodeTag(tree: Ast, i: Index) Node.Tag { + return tree.nodes.items(.tag)[i]; +} + +pub fn nodeToken(tree: Ast, i: Index) Index { + return tree.nodes.items(.main_token)[i]; +} + +pub fn nodeLHS(tree: Ast, i: Index) Index { + return tree.nodes.items(.lhs)[i]; +} + +pub fn nodeRHS(tree: Ast, i: Index) Index { + return tree.nodes.items(.rhs)[i]; +} + +pub const Index = u32; +pub const null_index: Index = 0; +pub const Node = struct { + tag: Tag, + main_token: Index, + lhs: Index = null_index, + rhs: Index = null_index, + + pub const Tag = enum { + /// an slice to extra field [LHS..RHS] + /// TOK : undefined + /// LHS : Index + /// RHS : Index + span, + + // ####### GlobalDecl ####### + + /// TOK : k_var + /// LHS : GlobalVarDecl + /// RHS : Expr? + global_variable, + + /// TOK : k_const + /// LHS : Type + /// RHS : Expr + global_constant, + + /// TOK : k_override + /// LHS : OverrideDecl + /// RHS : Expr + override, + + /// TOK : k_type + /// LHS : Type + /// RHS : -- + type_alias, + + /// TOK : k_const_assert + /// LHS : Expr + /// RHS : -- + const_assert, + + /// TOK : k_struct + /// LHS : span(struct_member) + /// RHS : -- + struct_decl, + /// TOK : ident + /// LHS : span(Attribute) + /// RHS : Type + struct_member, + + /// TOK : k_fn + /// LHS : FnProto + /// RHS : block + fn_decl, + /// TOK : ident + /// LHS : ? Attributes + /// RHS : type + fn_param, + + // ####### Statement ####### + + // block = span(Statement) + + /// TOK : k_return + /// LHS : Expr? + /// RHS : -- + @"return", + + /// TOK : k_discard + /// LHS : -- + /// RHS : -- + discard, + + /// TOK : k_loop + /// LHS : block + /// RHS : -- + loop, + + /// TOK : k_continuing + /// LHS : block + /// RHS : -- + continuing, + + /// TOK : k_break + /// LHS : Expr + /// RHS : -- + break_if, + + /// TOK : k_break + /// LHS : -- + /// RHS : -- + @"break", + + /// TOK : k_continue + /// LHS : -- + /// RHS : -- + @"continue", + + /// TOK : k_if + /// LHS : Expr + /// RHS : blcok + @"if", + /// RHS is else body + /// TOK : k_if + /// LHS : if + /// RHS : blcok + if_else, + /// TOK : k_if + /// LHS : if + /// RHS : if, if_else, if_else_if + if_else_if, + + /// TOK : k_switch + /// LHS : Expr + /// RHS : span(switch_case, switch_default, switch_case_default) + @"switch", + /// TOK : k_case + /// LHS : span(Expr) + /// RHS : block + switch_case, + /// TOK : k_default + /// LHS : block + /// RHS : -- + switch_default, + /// switch_case with default (`case 1, 2, default {}`) + /// TOK : k_case + /// LHS : span(Expr) + /// RHS : block + switch_case_default, + + /// TOK : k_var + /// LHS : VarDecl + /// RHS : Expr? + var_decl, + + /// TOK : k_const + /// LHS : Type? + /// RHS : Expr + const_decl, + + /// TOK : k_let + /// LHS : Type? + /// RHS : Expr + let_decl, + + /// TOK : k_while + /// LHS : Expr + /// RHS : block + @"while", + + /// TOK : k_for + /// LHS : ForHeader + /// RHS : block + @"for", + + /// TOK : plus_plus, minus_minus + /// LHS : Expr + increase_decrement, + + /// TOK : plus_equal, minus_equal, + /// times_equal, division_equal, + /// modulo_equal, and_equal, + /// or_equal, xor_equal, + /// shift_right_equal, shift_left_equal + /// LHS : Expr + /// RHS : Expr + compound_assign, + + /// TOK : equal + /// LHS : Expr + /// RHS : -- + phony_assign, + + // ####### Type ####### + + /// TOK : k_i32, k_u32, k_f32, k_f16, k_bool + /// LHS : -- + /// RHS : -- + scalar_type, + + /// TOK : k_sampler, k_comparison_sampler + /// LHS : -- + /// RHS : -- + sampler_type, + + /// TOK : k_vec2, k_vec3, k_vec4 + /// LHS : Type + /// RHS : -- + vector_type, + + /// TOK : k_mat2x2, k_mat2x3, k_mat2x4, + /// k_mat3x2, k_mat3x3, k_mat3x4, + /// k_mat4x2, k_mat4x3, k_mat4x4 + /// LHS : Type + /// RHS : -- + matrix_type, + + /// TOK : k_atomic + /// LHS : Type + /// RHS : -- + atomic_type, + + /// TOK : k_array + /// LHS : Type + /// RHS : Expr? + array_type, + + /// TOK : k_ptr + /// LHS : Type + /// RHS : PtrType + ptr_type, + + /// TOK : k_texture_1d, k_texture_2d, k_texture_2d_array, + /// k_texture_3d, k_texture_cube, k_texture_cube_array + /// LHS : Type + /// RHS : -- + sampled_texture_type, + + /// TOK : k_texture_multisampled_2d + /// LHS : Type + /// RHS : -- + multisampled_texture_type, + + /// TOK : k_texture_external + /// LHS : Type + /// RHS : -- + external_texture_type, + + /// TOK : k_texture_storage_1d, k_texture_storage_2d, + /// k_texture_storage_2d_array, k_texture_storage_3d + /// LHS : Index(Token(TexelFormat)) + /// RHS : Index(Token(AccessMode)) + storage_texture_type, + + /// TOK : k_texture_depth_2d, k_texture_depth_2d_array + /// k_texture_depth_cube, k_texture_depth_cube_array + /// k_texture_depth_multisampled_2d + /// LHS : -- + /// RHS : -- + depth_texture_type, + + /// TOK : ident + /// LHS : -- + /// RHS : -- + user_type, + + // ####### Attr ####### + + // TOK : attr + attr, + + /// TOK : attr + /// LHS : Expr + /// RHS : -- + attr_one_arg, + + /// TOK : attr + /// LHS : Index(Token(BuiltinValue)) + /// RHS : -- + attr_builtin, + + /// TOK : attr + /// LHS : WorkgroupSize + /// RHS : -- + attr_workgroup_size, + + /// TOK : attr + /// LHS : Index(Token(InterpolationType)) + /// RHS : Index(Token(InterpolationSample)) + attr_interpolate, + + // ####### Expr ####### + // see both Parser.zig and https://gpuweb.github.io/gpuweb/wgsl/#expression-grammar + + /// TOK : * + /// LHS : Expr + /// RHS : Expr + mul, + + /// TOK : / + /// LHS : Expr + /// RHS : Expr + div, + + /// TOK : % + /// LHS : Expr + /// RHS : Expr + mod, + + /// TOK : + + /// LHS : Expr + /// RHS : Expr + add, + + /// TOK : - + /// LHS : Expr + /// RHS : Expr + sub, + + /// TOK : << + /// LHS : Expr + /// RHS : Expr + shift_left, + + /// TOK : >> + /// LHS : Expr + /// RHS : Expr + shift_right, + + /// TOK : & + /// LHS : Expr + /// RHS : Expr + binary_and, + + /// TOK : | + /// LHS : Expr + /// RHS : Expr + binary_or, + + /// TOK : ^ + /// LHS : Expr + /// RHS : Expr + binary_xor, + + /// TOK : && + /// LHS : Expr + /// RHS : Expr + circuit_and, + + /// TOK : || + /// LHS : Expr + /// RHS : Expr + circuit_or, + + /// TOK : ! + /// LHS : Expr + /// RHS : -- + not, + + /// TOK : - + /// LHS : Expr + /// RHS : -- + negate, + + /// TOK : * + /// LHS : Expr + /// RHS : -- + deref, + + /// TOK : & + /// LHS : Expr + /// RHS : -- + addr_of, + + /// TOK : == + /// LHS : Expr + /// RHS : Expr + equal, + + /// TOK : != + /// LHS : Expr + /// RHS : Expr + not_equal, + + /// TOK : < + /// LHS : Expr + /// RHS : Expr + less, + + /// TOK : <= + /// LHS : Expr + /// RHS : Expr + less_equal, + + /// TOK : > + /// LHS : Expr + /// RHS : Expr + greater, + + /// TOK : >= + /// LHS : Expr + /// RHS : Expr + greater_equal, + + /// for identifier, array without element type specified, + /// vector prefix (e.g. vec2) and matrix prefix (e.g. mat2x2) LHS is null + /// see callExpr in Parser.zig if you don't understand this + /// + /// TOK : ident, k_array, 'scalar keywords', 'vector keywords', 'matrix keywords' + /// LHS : (scalar_type, vector_type, matrix_type, array_type)? + /// RHS : arguments (Expr span) + call, + + /// TOK : k_bitcast + /// LHS : Type + /// RHS : Expr + bitcast, + + /// TOK : ident + /// LHS : -- + /// RHS : -- + ident_expr, + + /// LHS is prefix expression + /// TOK : ident + /// LHS : Expr + component_access, + + /// LHS is prefix expression + /// TOK : bracket_left + /// LHS : Expr + /// RHS : Expr + index_access, + + // ####### Literals ####### + + /// TOK : k_true, k_false + /// LHS : -- + /// RHS : -- + bool_literal, + /// TOK : number + /// LHS : -- + /// RHS : -- + number_literal, + }; + + pub const GlobalVarDecl = struct { + /// span(Attr)? + attrs: Index = null_index, + /// Token(ident) + name: Index, + /// Token(AddrSpace)? + addr_space: Index = null_index, + /// Token(AccessMode)? + access_mode: Index = null_index, + /// Type? + type: Index = null_index, + }; + + pub const VarDecl = struct { + /// Token(ident) + name: Index, + /// Token(AddrSpace)? + addr_space: Index = null_index, + /// Token(AccessMode)? + access_mode: Index = null_index, + /// Type? + type: Index = null_index, + }; + + pub const OverrideDecl = struct { + /// span(Attr)? + attrs: Index = null_index, + /// Type? + type: Index = null_index, + }; + + pub const PtrType = struct { + /// Token(AddrSpace) + addr_space: Index, + /// Token(AccessMode)? + access_mode: Index = null_index, + }; + + pub const WorkgroupSize = struct { + /// Expr + x: Index, + /// Expr? + y: Index = null_index, + /// Expr? + z: Index = null_index, + }; + + pub const FnProto = struct { + /// span(Attr)? + attrs: Index = null_index, + /// span(fn_param)? + params: Index = null_index, + /// span(Attr)? + result_attrs: Index = null_index, + /// Type? + result_type: Index = null_index, + }; + + pub const IfStatement = struct { + /// Expr + cond: Index, + /// block + body: Index, + }; + + pub const ForHeader = struct { + /// var_decl, const_decl, let_decl, phony_assign, compound_assign + init: Index = null_index, + /// Expr + cond: Index = null_index, + /// call, phony_assign, compound_assign + update: Index = null_index, + }; +}; + +pub const BuiltinValue = enum { + vertex_index, + instance_index, + position, + front_facing, + frag_depth, + local_invocation_id, + local_invocation_index, + global_invocation_id, + workgroup_id, + num_workgroups, + sample_index, + sample_mask, +}; + +pub const InterpolationType = enum { + perspective, + linear, + flat, +}; + +pub const InterpolationSample = enum { + center, + centroid, + sample, +}; + +pub const AddressSpace = enum { + function, + private, + workgroup, + uniform, + storage, +}; + +pub const AccessMode = enum { + read, + write, + read_write, +}; + +pub const Attribute = enum { + invariant, + @"const", + vertex, + fragment, + compute, + @"align", + binding, + group, + id, + location, + size, + builtin, + workgroup_size, + interpolate, +}; + +pub const TexelFormat = enum { + rgba8unorm, + rgba8snorm, + rgba8uint, + rgba8sint, + rgba16uint, + rgba16sint, + rgba16float, + r32uint, + r32sint, + r32float, + rg32uint, + rg32sint, + rg32float, + rgba32uint, + rgba32sint, + rgba32float, + bgra8unorm, +}; diff --git a/libs/dusk/src/Parser.zig b/libs/dusk/src/Parser.zig new file mode 100644 index 00000000..ca718c91 --- /dev/null +++ b/libs/dusk/src/Parser.zig @@ -0,0 +1,1877 @@ +//! Based on cb76461a088a2b554f0248e7cf94d5a12b77e28e +const std = @import("std"); +const Ast = @import("Ast.zig"); +const Token = @import("Token.zig"); +const Tokenizer = @import("Tokenizer.zig"); +const Extension = @import("main.zig").Extension; +const ErrorMsg = @import("main.zig").ErrorMsg; +const comptimePrint = std.fmt.comptimePrint; +const fieldNames = std.meta.fieldNames; +const Parser = @This(); + +allocator: std.mem.Allocator, +source: [:0]const u8, +tok_i: Ast.Index, +tokens: Ast.TokenList.Slice, +nodes: std.MultiArrayList(Ast.Node), +extra: std.ArrayListUnmanaged(Ast.Index), +scratch: std.ArrayListUnmanaged(Ast.Index), +errors: std.ArrayListUnmanaged(ErrorMsg), +extensions: Extension.Array, + +pub fn deinit(p: *Parser) void { + p.nodes.deinit(p.allocator); + p.extra.deinit(p.allocator); + p.scratch.deinit(p.allocator); + for (p.errors.items) |*err_msg| err_msg.deinit(p.allocator); + p.errors.deinit(p.allocator); +} + +pub fn parseRoot(p: *Parser) !void { + const root = try p.addNode(.{ .tag = .span, .main_token = undefined }); + + while (try p.globalDirective()) |ext| { + p.extensions.set(ext, true); + } + + while (p.peekToken(.tag, 0) != .eof) { + const decl = try p.expectGlobalDeclRecoverable() orelse continue; + try p.scratch.append(p.allocator, decl); + } + + if (p.errors.items.len > 0) { + return error.Parsing; + } + + try p.extra.appendSlice(p.allocator, p.scratch.items); + p.nodes.items(.lhs)[root] = @intCast(Ast.Index, p.extra.items.len - p.scratch.items.len); + p.nodes.items(.rhs)[root] = @intCast(Ast.Index, p.extra.items.len); +} + +pub fn globalDirective(p: *Parser) !?Extension { + _ = p.eatToken(.k_enable) orelse return null; + const ext_token = try p.expectToken(.ident); + const ext = std.meta.stringToEnum(Extension, p.getToken(.loc, ext_token).slice(p.source)) orelse { + try p.addError(p.getToken(.loc, ext_token), "invalid extension", .{}, null); + return error.Parsing; + }; + return ext; +} + +pub fn expectGlobalDeclRecoverable(p: *Parser) !?Ast.Index { + return p.expectGlobalDecl() catch |err| switch (err) { + error.Parsing => { + p.findNextGlobal(); + return null; + }, + else => return err, + }; +} + +pub fn expectGlobalDecl(p: *Parser) !Ast.Index { + while (p.eatToken(.semicolon)) |_| {} + + const attrs = try p.attributeList(); + + if (try p.structDecl() orelse + try p.functionDecl(attrs)) |node| + { + return node; + } + + if (try p.globalConstDecl() orelse + try p.typeAliasDecl() orelse + try p.constAssert() orelse + try p.globalVarDecl(attrs) orelse + try p.globalOverrideDecl(attrs)) |node| + { + _ = try p.expectToken(.semicolon); + return node; + } + + try p.addError( + p.peekToken(.loc, 0), + "expected global declaration, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; +} + +pub fn attributeList(p: *Parser) !?Ast.Index { + const scratch_top = p.scratch.items.len; + defer p.scratch.shrinkRetainingCapacity(scratch_top); + while (true) { + const attr = try p.attribute() orelse break; + try p.scratch.append(p.allocator, attr); + } + const list = p.scratch.items[scratch_top..]; + if (list.len == 0) return null; + return try p.listToSpan(list); +} + +pub fn attribute(p: *Parser) !?Ast.Index { + const attr_token = p.eatToken(.attr) orelse return null; + const ident_tok = try p.expectToken(.ident); + const str = p.getToken(.loc, ident_tok).slice(p.source); + const tag = std.meta.stringToEnum(Ast.Attribute, str) orelse { + try p.addError( + p.getToken(.loc, ident_tok), + "unknown attribute '{s}'", + .{p.getToken(.loc, ident_tok).slice(p.source)}, + try ErrorMsg.Note.create( + p.allocator, + null, + "valid options are [{s}]", + .{fieldNames(Ast.Attribute)}, + ), + ); + return error.Parsing; + }; + var node = Ast.Node{ + .tag = undefined, + .main_token = attr_token, + }; + switch (tag) { + .invariant, + .@"const", + .vertex, + .fragment, + .compute, + => node.tag = .attr, + .@"align", + .binding, + .group, + .id, + .location, + .size, + .builtin, + => { + _ = try p.expectToken(.paren_left); + if (tag == .builtin) { + node.tag = .attr_builtin; + node.lhs = try p.expectBuiltinValue(); + } else { + node.tag = .attr_one_arg; + node.lhs = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected expression, but found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + } + _ = p.eatToken(.comma); + _ = try p.expectToken(.paren_right); + }, + .workgroup_size => { + _ = try p.expectToken(.paren_left); + + node.tag = .attr_workgroup_size; + var workgroup_size = Ast.Node.WorkgroupSize{ + .x = try p.expression() orelse { + try p.addError(p.peekToken(.loc, 0), "expected workgroup_size x parameter", .{}, null); + return error.Parsing; + }, + }; + + if (p.eatToken(.comma) != null and p.peekToken(.tag, 0) != .paren_right) { + workgroup_size.y = try p.expression() orelse { + try p.addError(p.peekToken(.loc, 0), "expected workgroup_size y parameter", .{}, null); + return error.Parsing; + }; + + if (p.eatToken(.comma) != null and p.peekToken(.tag, 0) != .paren_right) { + workgroup_size.z = try p.expression() orelse { + try p.addError(p.peekToken(.loc, 0), "expected workgroup_size z parameter", .{}, null); + return error.Parsing; + }; + + _ = p.eatToken(.comma); + } + } + + node.lhs = try p.addExtra(workgroup_size); + _ = try p.expectToken(.paren_right); + }, + .interpolate => { + _ = try p.expectToken(.paren_left); + + node.tag = .attr_interpolate; + node.lhs = try p.expectInterpolationType(); + + if (p.eatToken(.comma) != null and p.peekToken(.tag, 0) != .paren_right) { + node.rhs = try p.expectInterpolationSample(); + _ = p.eatToken(.comma); + _ = try p.expectToken(.paren_right); + } + + _ = try p.expectToken(.paren_right); + }, + } + + return try p.addNode(node); +} + +pub fn expectBuiltinValue(p: *Parser) !Ast.Index { + const token = p.advanceToken(); + if (p.getToken(.tag, token) == .ident) { + const str = p.getToken(.loc, token).slice(p.source); + if (std.meta.stringToEnum(Ast.BuiltinValue, str)) |_| return token; + } + + try p.addError( + p.getToken(.loc, token), + "unknown builtin value name '{s}'", + .{p.getToken(.loc, token).slice(p.source)}, + try ErrorMsg.Note.create( + p.allocator, + null, + "valid options are [{s}]", + .{fieldNames(Ast.BuiltinValue)}, + ), + ); + return error.Parsing; +} + +pub fn expectInterpolationType(p: *Parser) !Ast.Index { + const token = p.advanceToken(); + if (p.getToken(.tag, token) == .ident) { + const str = p.getToken(.loc, token).slice(p.source); + if (std.meta.stringToEnum(Ast.InterpolationType, str)) |_| return token; + } + + try p.addError( + p.getToken(.loc, token), + "unknown interpolation type name '{s}'", + .{p.getToken(.loc, token).slice(p.source)}, + try ErrorMsg.Note.create( + p.allocator, + null, + "valid options are [{s}]", + .{fieldNames(Ast.InterpolationType)}, + ), + ); + return error.Parsing; +} + +pub fn expectInterpolationSample(p: *Parser) !Ast.Index { + const token = p.advanceToken(); + if (p.getToken(.tag, token) == .ident) { + const str = p.getToken(.loc, token).slice(p.source); + if (std.meta.stringToEnum(Ast.InterpolationSample, str)) |_| return token; + } + + try p.addError( + p.getToken(.loc, token), + "unknown interpolation sample name '{s}'", + .{p.getToken(.loc, token).slice(p.source)}, + try ErrorMsg.Note.create( + p.allocator, + null, + "valid options are [{s}]", + .{fieldNames(Ast.InterpolationSample)}, + ), + ); + return error.Parsing; +} + +pub fn globalVarDecl(p: *Parser, attrs: ?Ast.Index) !?Ast.Index { + const var_token = p.eatToken(.k_var) orelse return null; + + // qualifier + var addr_space = Ast.null_index; + var access_mode = Ast.null_index; + if (p.eatToken(.less_than)) |_| { + addr_space = try p.expectAddressSpace(); + access_mode = if (p.eatToken(.comma)) |_| + try p.expectAccessMode() + else + Ast.null_index; + _ = try p.expectToken(.greater_than); + } + + // name, type + const name_token = try p.expectToken(.ident); + var var_type = Ast.null_index; + if (p.eatToken(.colon)) |_| { + var_type = try p.expectTypeSpecifier(); + } + + var initializer = Ast.null_index; + if (p.eatToken(.equal)) |_| { + initializer = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected initializer expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + } + + const extra = try p.addExtra(Ast.Node.GlobalVarDecl{ + .attrs = attrs orelse Ast.null_index, + .name = name_token, + .addr_space = addr_space, + .access_mode = access_mode, + .type = var_type, + }); + return try p.addNode(.{ + .tag = .global_variable, + .main_token = var_token, + .lhs = extra, + .rhs = initializer, + }); +} + +pub fn globalConstDecl(p: *Parser) !?Ast.Index { + const const_token = p.eatToken(.k_const) orelse return null; + + _ = try p.expectToken(.ident); + var const_type = Ast.null_index; + if (p.eatToken(.colon)) |_| { + const_type = try p.expectTypeSpecifier(); + } + + _ = try p.expectToken(.equal); + const initializer = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected initializer expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + + return try p.addNode(.{ + .tag = .global_constant, + .main_token = const_token, + .lhs = const_type, + .rhs = initializer, + }); +} + +pub fn globalOverrideDecl(p: *Parser, attrs: ?Ast.Index) !?Ast.Index { + const override_token = p.eatToken(.k_override) orelse return null; + + // name, type + _ = try p.expectToken(.ident); + var override_type = Ast.null_index; + if (p.eatToken(.colon)) |_| { + override_type = try p.expectTypeSpecifier(); + } + + var initializer = Ast.null_index; + if (p.eatToken(.equal)) |_| { + initializer = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected initializer expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + } + + const extra = try p.addExtra(Ast.Node.OverrideDecl{ + .attrs = attrs orelse Ast.null_index, + .type = override_type, + }); + return try p.addNode(.{ + .tag = .override, + .main_token = override_token, + .lhs = extra, + .rhs = initializer, + }); +} + +pub fn typeAliasDecl(p: *Parser) !?Ast.Index { + const type_token = p.eatToken(.k_type) orelse return null; + _ = try p.expectToken(.ident); + _ = try p.expectToken(.equal); + const value = try p.expectTypeSpecifier(); + return try p.addNode(.{ + .tag = .type_alias, + .main_token = type_token, + .lhs = value, + }); +} + +pub fn structDecl(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_struct) orelse return null; + _ = try p.expectToken(.ident); + _ = try p.expectToken(.brace_left); + + const scratch_top = p.scratch.items.len; + defer p.scratch.shrinkRetainingCapacity(scratch_top); + while (true) { + const attrs = try p.attributeList(); + const member = try p.structMember(attrs) orelse { + if (attrs != null) { + try p.addError( + p.peekToken(.loc, 0), + "expected struct member, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + } + break; + }; + try p.scratch.append(p.allocator, member); + _ = p.eatToken(.comma); + } + + _ = try p.expectToken(.brace_right); + + const list = p.scratch.items[scratch_top..]; + const members = try p.listToSpan(list); + + return try p.addNode(.{ + .tag = .struct_decl, + .main_token = main_token, + .lhs = members, + }); +} + +pub fn structMember(p: *Parser, attrs: ?Ast.Index) !?Ast.Index { + const name_token = p.eatToken(.ident) orelse return null; + _ = try p.expectToken(.colon); + const member_type = try p.expectTypeSpecifier(); + return try p.addNode(.{ + .tag = .struct_member, + .main_token = name_token, + .lhs = attrs orelse Ast.null_index, + .rhs = member_type, + }); +} + +pub fn constAssert(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_const_assert) orelse return null; + const expr = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + return try p.addNode(.{ + .tag = .const_assert, + .main_token = main_token, + .lhs = expr, + }); +} + +pub fn functionDecl(p: *Parser, attrs: ?Ast.Index) !?Ast.Index { + const fn_token = p.eatToken(.k_fn) orelse return null; + _ = try p.expectToken(.ident); + + _ = try p.expectToken(.paren_left); + const params = try p.parameterList() orelse Ast.null_index; + _ = try p.expectToken(.paren_right); + + var result_attrs = Ast.null_index; + var result_type = Ast.null_index; + if (p.eatToken(.arrow)) |_| { + result_attrs = try p.attributeList() orelse Ast.null_index; + result_type = try p.expectTypeSpecifier(); + } + + const body = try p.block() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected function body, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + + const fn_proto = try p.addExtra(Ast.Node.FnProto{ + .attrs = attrs orelse Ast.null_index, + .params = params, + .result_attrs = result_attrs, + .result_type = result_type, + }); + return try p.addNode(.{ + .tag = .fn_decl, + .main_token = fn_token, + .lhs = fn_proto, + .rhs = body, + }); +} + +pub fn parameterList(p: *Parser) !?Ast.Index { + const scratch_top = p.scratch.items.len; + defer p.scratch.shrinkRetainingCapacity(scratch_top); + while (true) { + const attrs = try p.attributeList(); + const param = try p.parameter(attrs) orelse { + if (attrs != null) { + try p.addError( + p.peekToken(.loc, 0), + "expected function parameter, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + } + break; + }; + try p.scratch.append(p.allocator, param); + if (p.eatToken(.comma) == null) break; + } + const list = p.scratch.items[scratch_top..]; + if (list.len == 0) return null; + return try p.listToSpan(list); +} + +pub fn parameter(p: *Parser, attrs: ?Ast.Index) !?Ast.Index { + const main_token = p.eatToken(.ident) orelse return null; + _ = try p.expectToken(.colon); + const param_type = try p.expectTypeSpecifier(); + return try p.addNode(.{ + .tag = .fn_param, + .main_token = main_token, + .lhs = attrs orelse Ast.null_index, + .rhs = param_type, + }); +} + +pub fn statementRecoverable(p: *Parser) !?Ast.Index { + while (true) { + return p.statement() catch |err| switch (err) { + error.Parsing => { + p.findNextStmt(); + switch (p.peekToken(.tag, 0)) { + .brace_right => return null, + .eof => return error.Parsing, + else => continue, + } + }, + else => return err, + }; + } +} + +/// for simplicity and better error messages, +/// we are putting all statements here +pub fn statement(p: *Parser) !?Ast.Index { + while (p.eatToken(.semicolon)) |_| {} + + if (try p.breakStatement() orelse + try p.breakIfStatement() orelse + try p.callExpr() orelse + try p.constAssert() orelse + try p.continueStatement() orelse + try p.discardStatement() orelse + try p.returnStatement() orelse + try p.varStatement() orelse + try p.varUpdateStatement()) |node| + { + _ = try p.expectToken(.semicolon); + return node; + } + + if (try p.block() orelse + try p.continuingStatement() orelse + try p.forStatement() orelse + try p.ifStatement() orelse + try p.loopStatement() orelse + try p.switchStatement() orelse + try p.whileStatement()) |node| + { + return node; + } + + return null; +} + +pub fn expectBlock(p: *Parser) error{ OutOfMemory, Parsing }!Ast.Index { + return try p.block() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected block statement, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; +} + +pub fn block(p: *Parser) error{ OutOfMemory, Parsing }!?Ast.Index { + _ = p.eatToken(.brace_left) orelse return null; + + const scratch_top = p.scratch.items.len; + defer p.scratch.shrinkRetainingCapacity(scratch_top); + + var failed = false; + while (true) { + const stmt = try p.statementRecoverable() orelse { + if (p.peekToken(.tag, 0) == .brace_right) break; + failed = true; + try p.addError( + p.peekToken(.loc, 0), + "expected statement, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + p.findNextStmt(); + continue; + }; + try p.scratch.append(p.allocator, stmt); + } + _ = try p.expectToken(.brace_right); + if (failed) return error.Parsing; + + const list = p.scratch.items[scratch_top..]; + return try p.listToSpan(list); +} + +pub fn breakStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_break) orelse return null; + return try p.addNode(.{ .tag = .@"break", .main_token = main_token }); +} + +pub fn breakIfStatement(p: *Parser) !?Ast.Index { + if (p.peekToken(.tag, 0) == .k_break and + p.peekToken(.tag, 1) == .k_if) + { + const main_token = p.advanceToken(); + _ = p.advanceToken(); + const cond = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected condition expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + return try p.addNode(.{ + .tag = .break_if, + .main_token = main_token, + .lhs = cond, + }); + } + return null; +} + +pub fn continueStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_continue) orelse return null; + return try p.addNode(.{ .tag = .@"continue", .main_token = main_token }); +} + +pub fn continuingStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_continuing) orelse return null; + const body = try p.expectBlock(); + return try p.addNode(.{ + .tag = .continuing, + .main_token = main_token, + .lhs = body, + }); +} + +pub fn discardStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_discard) orelse return null; + return try p.addNode(.{ .tag = .discard, .main_token = main_token }); +} + +pub fn forStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_for) orelse return null; + _ = try p.expectToken(.paren_left); + + // for init + const for_init = try p.callExpr() orelse + try p.varStatement() orelse + try p.varUpdateStatement() orelse + Ast.null_index; + _ = try p.expectToken(.semicolon); + + const for_cond = try p.expression() orelse Ast.null_index; + _ = try p.expectToken(.semicolon); + + // for update + const for_update = try p.callExpr() orelse + try p.varUpdateStatement() orelse + Ast.null_index; + + _ = try p.expectToken(.paren_right); + const body = try p.expectBlock(); + + const extra = try p.addExtra(Ast.Node.ForHeader{ + .init = for_init, + .cond = for_cond, + .update = for_update, + }); + return try p.addNode(.{ + .tag = .@"for", + .main_token = main_token, + .lhs = extra, + .rhs = body, + }); +} + +pub fn ifStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_if) orelse return null; + + const cond = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected condition expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + const body = try p.block() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected if body block, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + + if (p.eatToken(.k_else)) |_| { + const extra = try p.addExtra(Ast.Node.IfStatement{ + .cond = cond, + .body = body, + }); + + if (p.peekToken(.tag, 0) == .k_if) { + const else_if = try p.ifStatement() orelse unreachable; + return try p.addNode(.{ + .tag = .if_else_if, + .main_token = main_token, + .lhs = extra, + .rhs = else_if, + }); + } + + const else_body = try p.block() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected else body block, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + + return try p.addNode(.{ + .tag = .if_else, + .main_token = main_token, + .lhs = extra, + .rhs = else_body, + }); + } + + return try p.addNode(.{ + .tag = .@"if", + .main_token = main_token, + .lhs = cond, + .rhs = body, + }); +} + +pub fn loopStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_loop) orelse return null; + const body = try p.expectBlock(); + return try p.addNode(.{ + .tag = .loop, + .main_token = main_token, + .lhs = body, + }); +} + +pub fn returnStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_return) orelse return null; + const expr = try p.expression() orelse Ast.null_index; + return try p.addNode(.{ + .tag = .@"return", + .main_token = main_token, + .lhs = expr, + }); +} + +pub fn switchStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_switch) orelse return null; + + const expr = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected condition expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + + _ = try p.expectToken(.brace_left); + + const scratch_top = p.scratch.items.len; + defer p.scratch.shrinkRetainingCapacity(scratch_top); + while (true) { + if (p.eatToken(.k_default)) |default_token| { + _ = p.eatToken(.colon); + const default_body = try p.expectBlock(); + try p.scratch.append(p.allocator, try p.addNode(.{ + .tag = .switch_default, + .main_token = default_token, + .lhs = default_body, + })); + } else if (p.eatToken(.k_case)) |case_token| { + const cases_scratch_top = p.scratch.items.len; + + var has_default = false; + while (true) { + const case_expr = try p.expression() orelse { + if (p.eatToken(.k_default)) |_| continue; + break; + }; + _ = p.eatToken(.comma); + try p.scratch.append(p.allocator, case_expr); + } + const case_expr_list = p.scratch.items[cases_scratch_top..]; + + _ = p.eatToken(.colon); + const default_body = try p.expectBlock(); + + try p.scratch.append(p.allocator, try p.addNode(.{ + .tag = if (has_default) .switch_case_default else .switch_case, + .main_token = case_token, + .lhs = try p.listToSpan(case_expr_list), + .rhs = default_body, + })); + p.scratch.shrinkRetainingCapacity(cases_scratch_top); + } else { + break; + } + } + + _ = try p.expectToken(.brace_right); + + const case_list = p.scratch.items[scratch_top..]; + return try p.addNode(.{ + .tag = .@"switch", + .main_token = main_token, + .lhs = expr, + .rhs = try p.listToSpan(case_list), + }); +} + +pub fn varStatement(p: *Parser) !?Ast.Index { + if (p.eatToken(.k_var)) |var_token| { + var addr_space = Ast.null_index; + var access_mode = Ast.null_index; + if (p.eatToken(.less_than)) |_| { + addr_space = try p.expectAddressSpace(); + access_mode = if (p.eatToken(.comma)) |_| + try p.expectAccessMode() + else + Ast.null_index; + _ = try p.expectToken(.greater_than); + } + + const name_token = try p.expectToken(.ident); + var var_type = Ast.null_index; + if (p.eatToken(.colon)) |_| { + var_type = try p.expectTypeSpecifier(); + } + + var initializer = Ast.null_index; + if (p.eatToken(.equal)) |_| { + initializer = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected initializer expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + } + + const extra = try p.addExtra(Ast.Node.VarDecl{ + .name = name_token, + .addr_space = addr_space, + .access_mode = access_mode, + .type = var_type, + }); + return try p.addNode(.{ + .tag = .var_decl, + .main_token = var_token, + .lhs = extra, + .rhs = initializer, + }); + } + + const const_let_token = p.eatToken(.k_const) orelse p.eatToken(.k_let) orelse return null; + _ = try p.expectToken(.ident); + var const_let_type = Ast.null_index; + if (p.eatToken(.colon)) |_| { + const_let_type = try p.expectTypeSpecifier(); + } + + _ = try p.expectToken(.equal); + const initializer = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected initializer expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + + return try p.addNode(.{ + .tag = if (p.getToken(.tag, const_let_token) == .k_const) + .const_decl + else + .let_decl, + .main_token = const_let_token, + .lhs = const_let_type, + .rhs = initializer, + }); +} + +pub fn varUpdateStatement(p: *Parser) !?Ast.Index { + if (p.eatToken(.underscore)) |_| { + const equal_token = try p.expectToken(.equal); + const expr = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + return try p.addNode(.{ + .tag = .phony_assign, + .main_token = equal_token, + .lhs = expr, + }); + } else if (try p.lhsExpression()) |lhs| { + const op_token = p.advanceToken(); + switch (p.getToken(.tag, op_token)) { + .plus_plus, .minus_minus => { + return try p.addNode(.{ + .tag = .increase_decrement, + .main_token = op_token, + .lhs = lhs, + }); + }, + .equal, + .plus_equal, + .minus_equal, + .times_equal, + .division_equal, + .modulo_equal, + .and_equal, + .or_equal, + .xor_equal, + .shift_right_equal, + .shift_left_equal, + => { + const expr = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + return try p.addNode(.{ + .tag = .compound_assign, + .main_token = op_token, + .lhs = lhs, + .rhs = expr, + }); + }, + else => { + try p.addError( + p.getToken(.loc, op_token), + "invalid assignment operator '{s}'", + .{p.getToken(.tag, op_token).symbol()}, + null, + ); + return error.Parsing; + }, + } + } + + return null; +} + +pub fn whileStatement(p: *Parser) !?Ast.Index { + const main_token = p.eatToken(.k_while) orelse return null; + const cond = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected condition expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + const body = try p.expectBlock(); + return try p.addNode(.{ + .tag = .@"while", + .main_token = main_token, + .lhs = cond, + .rhs = body, + }); +} + +pub fn expectTypeSpecifier(p: *Parser) error{ OutOfMemory, Parsing }!Ast.Index { + return try p.typeSpecifier() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected type sepecifier, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; +} + +pub fn typeSpecifier(p: *Parser) !?Ast.Index { + if (p.peekToken(.tag, 0) == .ident) { + const main_token = p.advanceToken(); + return try p.addNode(.{ .tag = .user_type, .main_token = main_token }); + } + return p.typeSpecifierWithoutIdent(); +} + +pub fn typeSpecifierWithoutIdent(p: *Parser) !?Ast.Index { + if (p.isVectorPrefix() or p.isMatrixPrefix()) { + const main_token = p.advanceToken(); + _ = try p.expectToken(.less_than); + const elem_type = try p.expectTypeSpecifier(); + _ = try p.expectToken(.greater_than); + return try p.addNode(.{ + .tag = if (p.isVectorPrefix()) .vector_type else .matrix_type, + .main_token = main_token, + .lhs = elem_type, + }); + } + + const main_token = p.tok_i; + switch (p.getToken(.tag, main_token)) { + .k_i32, + .k_u32, + .k_f32, + .k_f16, + .k_bool, + => { + _ = p.advanceToken(); + return try p.addNode(.{ .tag = .scalar_type, .main_token = main_token }); + }, + .k_sampler, .k_comparison_sampler => { + _ = p.advanceToken(); + return try p.addNode(.{ .tag = .sampler_type, .main_token = main_token }); + }, + .k_atomic => { + _ = p.advanceToken(); + _ = try p.expectToken(.less_than); + const elem_type = try p.expectTypeSpecifier(); + _ = try p.expectToken(.greater_than); + return try p.addNode(.{ + .tag = .atomic_type, + .main_token = main_token, + .lhs = elem_type, + }); + }, + .k_array => { + _ = p.advanceToken(); + _ = try p.expectToken(.less_than); + const elem_type = try p.expectTypeSpecifier(); + var size = Ast.null_index; + if (p.eatToken(.comma)) |_| { + size = try p.elementCountExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected array size expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + } + _ = try p.expectToken(.greater_than); + return try p.addNode(.{ + .tag = .array_type, + .main_token = main_token, + .lhs = elem_type, + .rhs = size, + }); + }, + .k_ptr => { + _ = p.advanceToken(); + _ = try p.expectToken(.less_than); + + const addr_space = try p.expectAddressSpace(); + _ = try p.expectToken(.comma); + const elem_type = try p.expectTypeSpecifier(); + var access_mode = Ast.null_index; + if (p.eatToken(.comma)) |_| { + access_mode = try p.expectAccessMode(); + } + _ = try p.expectToken(.greater_than); + + const extra = try p.addExtra(Ast.Node.PtrType{ + .addr_space = addr_space, + .access_mode = access_mode, + }); + return try p.addNode(.{ + .tag = .ptr_type, + .main_token = main_token, + .lhs = elem_type, + .rhs = extra, + }); + }, + .k_texture_sampled_1d, + .k_texture_sampled_2d, + .k_texture_sampled_2d_array, + .k_texture_sampled_3d, + .k_texture_sampled_cube, + .k_texture_sampled_cube_array, + => { + _ = p.advanceToken(); + _ = try p.expectToken(.less_than); + const elem_type = try p.expectTypeSpecifier(); + _ = try p.expectToken(.greater_than); + return try p.addNode(.{ + .tag = .sampled_texture_type, + .main_token = main_token, + .lhs = elem_type, + }); + }, + .k_texture_multisampled_2d => { + return try p.addNode(.{ + .tag = .multisampled_texture_type, + .main_token = main_token, + }); + }, + .k_texture_external => { + return try p.addNode(.{ + .tag = .external_texture_type, + .main_token = main_token, + }); + }, + .k_texture_depth_2d, + .k_texture_depth_2d_array, + .k_texture_depth_cube, + .k_texture_depth_cube_array, + .k_texture_depth_multisampled_2d, + => { + return try p.addNode(.{ + .tag = .depth_texture_type, + .main_token = main_token, + }); + }, + .k_texture_storage_1d, + .k_texture_storage_2d, + .k_texture_storage_2d_array, + .k_texture_storage_3d, + => { + _ = p.advanceToken(); + _ = try p.expectToken(.less_than); + const texel_format = try p.expectTexelFormat(); + _ = try p.expectToken(.comma); + const access_mode = try p.expectAccessMode(); + _ = try p.expectToken(.greater_than); + return try p.addNode(.{ + .tag = .storage_texture_type, + .main_token = main_token, + .lhs = texel_format, + .rhs = access_mode, + }); + }, + else => return null, + } +} + +pub fn isVectorPrefix(p: *Parser) bool { + return switch (p.peekToken(.tag, 0)) { + .k_vec2, + .k_vec3, + .k_vec4, + => true, + else => false, + }; +} + +pub fn isMatrixPrefix(p: *Parser) bool { + return switch (p.peekToken(.tag, 0)) { + .k_mat2x2, + .k_mat2x3, + .k_mat2x4, + .k_mat3x2, + .k_mat3x3, + .k_mat3x4, + .k_mat4x2, + .k_mat4x3, + .k_mat4x4, + => true, + else => false, + }; +} + +pub fn expectAddressSpace(p: *Parser) !Ast.Index { + const token = p.advanceToken(); + if (p.getToken(.tag, token) == .ident) { + const str = p.getToken(.loc, token).slice(p.source); + if (std.meta.stringToEnum(Ast.AddressSpace, str)) |_| return token; + } + + try p.addError( + p.getToken(.loc, token), + "unknown address space '{s}'", + .{p.getToken(.loc, token).slice(p.source)}, + try ErrorMsg.Note.create( + p.allocator, + null, + "valid options are [{s}]", + .{fieldNames(Ast.AddressSpace)}, + ), + ); + return error.Parsing; +} + +pub fn expectAccessMode(p: *Parser) !Ast.Index { + const token = p.advanceToken(); + if (p.getToken(.tag, token) == .ident) { + const str = p.getToken(.loc, token).slice(p.source); + if (std.meta.stringToEnum(Ast.AccessMode, str)) |_| return token; + } + + try p.addError( + p.getToken(.loc, token), + "unknown access mode '{s}'", + .{p.getToken(.loc, token).slice(p.source)}, + try ErrorMsg.Note.create( + p.allocator, + null, + "valid options are [{s}]", + .{fieldNames(Ast.AccessMode)}, + ), + ); + return error.Parsing; +} + +pub fn expectTexelFormat(p: *Parser) !Ast.Index { + const token = p.advanceToken(); + if (p.getToken(.tag, token) == .ident) { + const str = p.getToken(.loc, token).slice(p.source); + if (std.meta.stringToEnum(Ast.TexelFormat, str)) |_| return token; + } + + try p.addError( + p.getToken(.loc, token), + "unknown address space '{s}'", + .{p.getToken(.loc, token).slice(p.source)}, + try ErrorMsg.Note.create( + p.allocator, + null, + "valid options are [{s}]", + .{fieldNames(Ast.TexelFormat)}, + ), + ); + return error.Parsing; +} + +pub fn expectParenExpr(p: *Parser) !Ast.Index { + _ = try p.expectToken(.paren_left); + const expr = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse expression '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + _ = try p.expectToken(.paren_right); + return expr; +} + +pub fn callExpr(p: *Parser) !?Ast.Index { + const main_token = p.tok_i; + var lhs = Ast.null_index; + + // function call + if (p.peekToken(.tag, 0) == .ident and p.peekToken(.tag, 1) == .paren_left) { + _ = p.advanceToken(); + } + // without template args ('vec2', 'array', etc) + else if (p.peekToken(.tag, 1) != .less_than and + (p.isVectorPrefix() or + p.isMatrixPrefix() or + p.peekToken(.tag, 0) == .k_array)) + { + _ = p.advanceToken(); + } else { + // maybe with template args ('i32', 'vec2', 'array', etc) + const type_node = try p.typeSpecifierWithoutIdent() orelse return null; + const tag = p.nodes.items(.tag)[type_node]; + switch (tag) { + .scalar_type, + .vector_type, + .matrix_type, + .array_type, + => lhs = type_node, + else => { + try p.addError( + p.getToken(.loc, main_token), + "type '{s}' can not be constructed", + .{p.getToken(.tag, main_token).symbol()}, + null, + ); + return error.Parsing; + }, + } + } + + const rhs = try p.expectArgumentListExpr(); + return try p.addNode(.{ + .tag = .call, + .main_token = main_token, + .lhs = lhs, + .rhs = rhs, + }); +} + +pub fn expectArgumentListExpr(p: *Parser) !Ast.Index { + _ = try p.expectToken(.paren_left); + + const scratch_top = p.scratch.items.len; + defer p.scratch.shrinkRetainingCapacity(scratch_top); + while (true) { + const expr = try p.expression() orelse break; + try p.scratch.append(p.allocator, expr); + if (p.eatToken(.comma) == null) break; + } + + _ = try p.expectToken(.paren_right); + + const list = p.scratch.items[scratch_top..]; + return p.listToSpan(list); +} + +pub fn expression(p: *Parser) !?Ast.Index { + const lhs_unary = try p.unaryExpr() orelse return null; + if (try p.bitwiseExpr(lhs_unary)) |bitwise| return bitwise; + const lhs = try p.expectRelationalExpr(lhs_unary); + return try p.expectShortCircuitExpr(lhs); +} + +pub fn lhsExpression(p: *Parser) !?Ast.Index { + if (p.eatToken(.ident)) |ident_token| { + return try p.componentOrSwizzleSpecifier( + try p.addNode(.{ .tag = .ident_expr, .main_token = ident_token }), + ); + } + + if (p.eatToken(.paren_left)) |_| { + const expr = try p.lhsExpression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected lhs expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + _ = try p.expectToken(.paren_right); + return try p.componentOrSwizzleSpecifier(expr); + } + + if (p.eatToken(.star)) |star_token| { + return try p.addNode(.{ + .tag = .deref, + .main_token = star_token, + .lhs = try p.lhsExpression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected lhs expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }, + }); + } + + if (p.eatToken(.@"and")) |addr_of_token| { + return try p.addNode(.{ + .tag = .addr_of, + .main_token = addr_of_token, + .lhs = try p.lhsExpression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected lhs expression, found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }, + }); + } + + return null; +} + +pub fn singularExpr(p: *Parser) !?Ast.Index { + const prefix = try p.primaryExpr() orelse return null; + return try p.componentOrSwizzleSpecifier(prefix); +} + +pub fn primaryExpr(p: *Parser) !?Ast.Index { + const main_token = p.tok_i; + if (try p.callExpr()) |call| return call; + switch (p.getToken(.tag, main_token)) { + .k_true, .k_false => { + _ = p.advanceToken(); + return try p.addNode(.{ .tag = .bool_literal, .main_token = main_token }); + }, + .number => { + _ = p.advanceToken(); + return try p.addNode(.{ .tag = .number_literal, .main_token = main_token }); + }, + .k_bitcast => { + _ = p.advanceToken(); + _ = try p.expectToken(.less_than); + const dest_type = try p.expectTypeSpecifier(); + _ = try p.expectToken(.greater_than); + const expr = try p.expectParenExpr(); + return try p.addNode(.{ + .tag = .bitcast, + .main_token = main_token, + .lhs = dest_type, + .rhs = expr, + }); + }, + .paren_left => return try p.expectParenExpr(), + .ident => { + _ = p.advanceToken(); + return try p.addNode(.{ .tag = .ident_expr, .main_token = main_token }); + }, + else => { + return null; + }, + } +} + +pub fn elementCountExpr(p: *Parser) !?Ast.Index { + const left = try p.unaryExpr() orelse return null; + if (try p.bitwiseExpr(left)) |right| return right; + return try p.expectMathExpr(left); +} + +pub fn unaryExpr(p: *Parser) error{ OutOfMemory, Parsing }!?Ast.Index { + const op_token = p.tok_i; + const op: Ast.Node.Tag = switch (p.getToken(.tag, op_token)) { + .bang, .tilde => .not, + .minus => .negate, + .star => .deref, + .@"and" => .addr_of, + else => return p.singularExpr(), + }; + _ = p.advanceToken(); + + const expr = try p.unaryExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse right side of '{s}' expression", + .{p.getToken(.tag, op_token).symbol()}, + null, + ); + return error.Parsing; + }; + + return try p.addNode(.{ + .tag = op, + .main_token = op_token, + .lhs = expr, + }); +} + +pub fn expectRelationalExpr(p: *Parser, lhs_unary: Ast.Index) !Ast.Index { + const lhs = try p.expectShiftExpr(lhs_unary); + const op_token = p.tok_i; + const op: Ast.Node.Tag = switch (p.getToken(.tag, op_token)) { + .equal_equal => .equal, + .not_equal => .not_equal, + .less_than => .less, + .less_than_equal => .less_equal, + .greater_than => .greater, + .greater_than_equal => .greater_equal, + else => return lhs, + }; + _ = p.advanceToken(); + + const rhs_unary = try p.unaryExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse right side of '{s}' expression", + .{p.getToken(.tag, op_token).symbol()}, + null, + ); + return error.Parsing; + }; + const rhs = try p.expectShiftExpr(rhs_unary); + return try p.addNode(.{ + .tag = op, + .main_token = op_token, + .lhs = lhs, + .rhs = rhs, + }); +} + +pub fn expectShortCircuitExpr(p: *Parser, lhs_relational: Ast.Index) !Ast.Index { + var lhs = lhs_relational; + + const op_token = p.tok_i; + const op: Ast.Node.Tag = switch (p.getToken(.tag, op_token)) { + .and_and => .circuit_and, + .or_or => .circuit_or, + else => return lhs, + }; + + while (p.peekToken(.tag, 0) == p.getToken(.tag, op_token)) { + _ = p.advanceToken(); + + const rhs_unary = try p.unaryExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse right side of '{s}' expression", + .{p.getToken(.tag, op_token).symbol()}, + null, + ); + return error.Parsing; + }; + const rhs = try p.expectRelationalExpr(rhs_unary); + + lhs = try p.addNode(.{ + .tag = op, + .main_token = op_token, + .lhs = lhs, + .rhs = rhs, + }); + } + + return lhs; +} + +pub fn bitwiseExpr(p: *Parser, lhs: Ast.Index) !?Ast.Index { + const op_token = p.tok_i; + const op: Ast.Node.Tag = switch (p.getToken(.tag, op_token)) { + .@"and" => .binary_and, + .@"or" => .binary_or, + .xor => .binary_xor, + else => return null, + }; + _ = p.advanceToken(); + + var lhs_result = lhs; + while (true) { + const rhs = try p.unaryExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse right side of '{s}' expression", + .{p.getToken(.tag, op_token).symbol()}, + null, + ); + return error.Parsing; + }; + + lhs_result = try p.addNode(.{ + .tag = op, + .main_token = op_token, + .lhs = lhs_result, + .rhs = rhs, + }); + + if (p.peekToken(.tag, 0) != p.getToken(.tag, op_token)) return lhs_result; + } +} + +pub fn expectShiftExpr(p: *Parser, lhs: Ast.Index) !Ast.Index { + const op_token = p.tok_i; + const op: Ast.Node.Tag = switch (p.getToken(.tag, op_token)) { + .shift_left => .shift_left, + .shift_right => .shift_right, + else => return try p.expectMathExpr(lhs), + }; + _ = p.advanceToken(); + + const rhs = try p.unaryExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse right side of '{s}' expression", + .{p.getToken(.tag, op_token).symbol()}, + null, + ); + return error.Parsing; + }; + + return try p.addNode(.{ + .tag = op, + .main_token = op_token, + .lhs = lhs, + .rhs = rhs, + }); +} + +pub fn expectMathExpr(p: *Parser, left: Ast.Index) !Ast.Index { + const right = try p.expectMultiplicativeExpr(left); + return p.expectAdditiveExpr(right); +} + +pub fn expectAdditiveExpr(p: *Parser, lhs_mul: Ast.Index) !Ast.Index { + var lhs = lhs_mul; + while (true) { + const op_token = p.tok_i; + const op: Ast.Node.Tag = switch (p.getToken(.tag, op_token)) { + .plus => .add, + .minus => .sub, + else => return lhs, + }; + _ = p.advanceToken(); + const unary = try p.unaryExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse right side of '{s}' expression", + .{p.getToken(.tag, op_token).symbol()}, + null, + ); + return error.Parsing; + }; + const rhs = try p.expectMultiplicativeExpr(unary); + lhs = try p.addNode(.{ + .tag = op, + .main_token = op_token, + .lhs = lhs, + .rhs = rhs, + }); + } +} + +pub fn expectMultiplicativeExpr(p: *Parser, lhs_unary: Ast.Index) !Ast.Index { + var lhs = lhs_unary; + while (true) { + const op_token = p.tok_i; + const node_tag: Ast.Node.Tag = switch (p.peekToken(.tag, 0)) { + .star => .mul, + .division => .div, + .mod => .mod, + else => return lhs, + }; + _ = p.advanceToken(); + const rhs = try p.unaryExpr() orelse { + try p.addError( + p.peekToken(.loc, 0), + "unable to parse right side of '{s}' expression", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + lhs = try p.addNode(.{ + .tag = node_tag, + .main_token = op_token, + .lhs = lhs, + .rhs = rhs, + }); + } +} + +pub fn componentOrSwizzleSpecifier(p: *Parser, prefix: Ast.Index) !Ast.Index { + var prefix_result = prefix; + while (true) { + if (p.eatToken(.period)) |_| { + const member_token = try p.expectToken(.ident); + prefix_result = try p.addNode(.{ + .tag = .component_access, + .main_token = member_token, + .lhs = prefix_result, + }); + } else if (p.eatToken(.bracket_left)) |bracket_left_token| { + const index_expr = try p.expression() orelse { + try p.addError( + p.peekToken(.loc, 0), + "expected expression, but found '{s}'", + .{p.peekToken(.tag, 0).symbol()}, + null, + ); + return error.Parsing; + }; + _ = try p.expectToken(.bracket_right); + prefix_result = try p.addNode(.{ + .tag = .index_access, + .main_token = bracket_left_token, + .lhs = prefix_result, + .rhs = index_expr, + }); + } else return prefix_result; + } +} + +fn findNextGlobal(p: *Parser) void { + var level: Ast.Index = 0; + while (true) { + switch (p.peekToken(.tag, 0)) { + .k_fn, + .k_var, + .k_const, + .k_override, + .k_struct, + .attr, + => { + if (level == 0) return; + }, + .semicolon => { + if (level == 0) { + _ = p.advanceToken(); + return; + } + }, + .brace_left, + .bracket_left, + .paren_left, + => { + level += 1; + }, + .brace_right => { + if (level == 0) { + _ = p.advanceToken(); + return; + } + level -= 1; + }, + .bracket_right, .paren_right => { + if (level != 0) level -= 1; + }, + .eof => return, + else => {}, + } + _ = p.advanceToken(); + } +} + +fn findNextStmt(p: *Parser) void { + var level: Ast.Index = 0; + while (true) { + switch (p.peekToken(.tag, 0)) { + .semicolon => { + if (level == 0) { + _ = p.advanceToken(); + return; + } + }, + .brace_left => { + level += 1; + }, + .brace_right => { + if (level == 0) { + _ = p.advanceToken(); + return; + } + level -= 1; + }, + .eof => return, + else => {}, + } + _ = p.advanceToken(); + } +} + +pub fn addError( + p: *Parser, + loc: Token.Loc, + comptime format: []const u8, + args: anytype, + note: ?ErrorMsg.Note, +) !void { + const err_msg = try ErrorMsg.create(p.allocator, loc, format, args, note); + try p.errors.append(p.allocator, err_msg); +} + +fn listToSpan(p: *Parser, list: []const Ast.Index) !Ast.Index { + try p.extra.appendSlice(p.allocator, list); + return p.addNode(.{ + .tag = .span, + .main_token = undefined, + .lhs = @intCast(Ast.Index, p.extra.items.len - list.len), + .rhs = @intCast(Ast.Index, p.extra.items.len), + }); +} + +pub fn addNode(p: *Parser, node: Ast.Node) error{OutOfMemory}!Ast.Index { + const i = @intCast(Ast.Index, p.nodes.len); + try p.nodes.append(p.allocator, node); + return i; +} + +fn addExtra(p: *Parser, extra: anytype) error{OutOfMemory}!Ast.Index { + const fields = std.meta.fields(@TypeOf(extra)); + try p.extra.ensureUnusedCapacity(p.allocator, fields.len); + const result = @intCast(Ast.Index, p.extra.items.len); + inline for (fields) |field| { + comptime std.debug.assert(field.type == Ast.Index or field.type == Ast.Index); + p.extra.appendAssumeCapacity(@field(extra, field.name)); + } + return result; +} + +pub fn getToken( + p: Parser, + comptime field: Ast.TokenList.Field, + index: Ast.Index, +) std.meta.fieldInfo(Token, field).type { + return p.tokens.items(field)[index]; +} + +pub fn peekToken( + p: Parser, + comptime field: Ast.TokenList.Field, + offset: Ast.Index, +) std.meta.fieldInfo(Token, field).type { + return p.getToken(field, p.tok_i + offset); +} + +pub fn advanceToken(p: *Parser) Ast.Index { + const prev = p.tok_i; + p.tok_i = std.math.min(prev +| 1, p.tokens.len); + return prev; +} + +pub fn eatToken(p: *Parser, tag: Token.Tag) ?Ast.Index { + return if (p.peekToken(.tag, 0) == tag) p.advanceToken() else null; +} + +pub fn expectToken(p: *Parser, tag: Token.Tag) !Ast.Index { + const token = p.advanceToken(); + if (p.getToken(.tag, token) == tag) return token; + + try p.addError( + p.getToken(.loc, token), + "expected '{s}', but found '{s}'", + .{ tag.symbol(), p.getToken(.tag, token).symbol() }, + null, + ); + return error.Parsing; +} diff --git a/libs/dusk/src/Token.zig b/libs/dusk/src/Token.zig new file mode 100644 index 00000000..e563b585 --- /dev/null +++ b/libs/dusk/src/Token.zig @@ -0,0 +1,467 @@ +const std = @import("std"); + +tag: Tag, +loc: Loc, + +pub const Loc = struct { + start: u32, + end: u32, + + pub const Extra = struct { + line: u32, + col: u32, + line_start: u32, + line_end: u32, + }; + + pub fn slice(self: Loc, source: []const u8) []const u8 { + return source[self.start..self.end]; + } + + pub fn extraInfo(self: Loc, source: []const u8) Extra { + var result = Extra{ + .line = 1, + .col = 1, + .line_start = 0, + .line_end = @intCast(u32, source.len), + }; + + for (source[0..self.start], 0..) |c, i| { + if (c == '\n') { + result.line += 1; + result.line_start = @intCast(u32, i) + 1; + } + } + + for (source[self.end..], 0..) |c, i| { + if (c == '\n') { + result.line_end = self.end + @intCast(u32, i); + break; + } + } + + result.col += self.start - result.line_start; + return result; + } +}; + +pub const Tag = enum { + eof, + invalid, + + ident, + /// any number literal + number, + + /// '&' + @"and", + /// '&&' + and_and, + /// '->' + arrow, + /// '@' + attr, + /// '/' + division, + /// '!' + bang, + /// '{' + brace_left, + /// '}' + brace_right, + /// '[' + bracket_left, + /// ']' + bracket_right, + /// ':' + colon, + /// ',' + comma, + /// '=' + equal, + /// '==' + equal_equal, + /// '>' + greater_than, + /// '>=' + greater_than_equal, + /// '>>' + shift_right, + /// '<' + less_than, + /// '<=' + less_than_equal, + /// '<<' + shift_left, + /// '%' + mod, + /// '-' + minus, + /// '--' + minus_minus, + /// '!=' + not_equal, + /// '.' + period, + /// '+' + plus, + /// '++' + plus_plus, + /// '|' + @"or", + /// '||' + or_or, + /// '(' + paren_left, + /// ')' + paren_right, + /// ';' + semicolon, + /// '*' + star, + /// '~' + tilde, + /// '_' + underscore, + /// '^' + xor, + /// '+=' + plus_equal, + /// '-=' + minus_equal, + /// '*=' + times_equal, + /// '/=' + division_equal, + /// '%=' + modulo_equal, + /// '&=' + and_equal, + /// '|=' + or_equal, + /// '^=' + xor_equal, + /// '>>=' + shift_right_equal, + /// '<<=' + shift_left_equal, + + /// 'array' + k_array, + /// 'atomic' + k_atomic, + /// 'bitcast' + k_bitcast, + /// 'bool' + k_bool, + /// 'break' + k_break, + /// 'case' + k_case, + /// 'const' + k_const, + /// 'continue' + k_continue, + /// 'continuing' + k_continuing, + /// 'discard' + k_discard, + /// 'default' + k_default, + /// 'else' + k_else, + /// 'enable' + k_enable, + /// 'f16' + k_f16, + /// 'f32' + k_f32, + /// 'fallthrough' + k_fallthrough, + /// 'false' + k_false, + /// 'fn' + k_fn, + /// 'for' + k_for, + /// 'i32' + k_i32, + /// 'if' + k_if, + /// 'let' + k_let, + /// 'loop' + k_loop, + /// 'mat2x2' + k_mat2x2, + /// 'mat2x3' + k_mat2x3, + /// 'mat2x4' + k_mat2x4, + /// 'mat3x2' + k_mat3x2, + /// 'mat3x3' + k_mat3x3, + /// 'mat3x4' + k_mat3x4, + /// 'mat4x2' + k_mat4x2, + /// 'mat4x3' + k_mat4x3, + /// 'mat4x4' + k_mat4x4, + /// 'override' + k_override, + /// 'ptr' + k_ptr, + /// 'return' + k_return, + /// 'sampler' + k_sampler, + /// 'sampler_comparison' + k_comparison_sampler, + /// 'const_assert' + k_const_assert, + /// 'struct' + k_struct, + /// 'switch' + k_switch, + /// 'texture_depth_2d' + k_texture_depth_2d, + /// 'texture_depth_2d_array' + k_texture_depth_2d_array, + /// 'texture_depth_cube' + k_texture_depth_cube, + /// 'texture_depth_cube_array' + k_texture_depth_cube_array, + /// 'texture_depth_multisampled_2d' + k_texture_depth_multisampled_2d, + /// 'texture_external' + k_texture_external, + /// 'texture_multisampled_2d' + k_texture_multisampled_2d, + /// 'texture_1d' + k_texture_sampled_1d, + /// 'texture_2d' + k_texture_sampled_2d, + /// 'texture_2d_array' + k_texture_sampled_2d_array, + /// 'texture_3d' + k_texture_sampled_3d, + /// 'texture_cube' + k_texture_sampled_cube, + /// 'texture_cube_array' + k_texture_sampled_cube_array, + /// 'texture_storage_1d' + k_texture_storage_1d, + /// 'texture_storage_2d' + k_texture_storage_2d, + /// 'texture_storage_2d_array' + k_texture_storage_2d_array, + /// 'texture_storage_3d' + k_texture_storage_3d, + /// 'true' + k_true, + /// 'type' + k_type, + /// 'u32' + k_u32, + /// 'var' + k_var, + /// 'vec2' + k_vec2, + /// 'vec3' + k_vec3, + /// 'vec4' + k_vec4, + /// 'while' + k_while, + + pub fn symbol(self: Tag) []const u8 { + return switch (self) { + .eof => "EOF", + .invalid => "invalid bytes", + .ident => "an identifier", + .number => "a number literal", + .@"and" => "&", + .and_and => "&&", + .arrow => "->", + .attr => "@", + .division => "/", + .bang => "!", + .brace_left => "{", + .brace_right => "}", + .bracket_left => "[", + .bracket_right => "]", + .colon => ":", + .comma => ",", + .equal => "=", + .equal_equal => "==", + .greater_than => ">", + .greater_than_equal => ">=", + .shift_right => ">>", + .less_than => "<", + .less_than_equal => "<=", + .shift_left => "<<", + .mod => "%", + .minus => "-", + .minus_minus => "--", + .not_equal => "!=", + .period => ".", + .plus => "+", + .plus_plus => "++", + .@"or" => "|", + .or_or => "||", + .paren_left => "(", + .paren_right => ")", + .semicolon => ";", + .star => "*", + .tilde => "~", + .underscore => "_", + .xor => "^", + .plus_equal => "+=", + .minus_equal => "-=", + .times_equal => "*=", + .division_equal => "/=", + .modulo_equal => "%=", + .and_equal => "&=", + .or_equal => "|=", + .xor_equal => "^=", + .shift_right_equal => ">>=", + .shift_left_equal => "<<=", + .k_array => "array", + .k_atomic => "atomic", + .k_bitcast => "bitcast", + .k_bool => "bool", + .k_break => "break", + .k_case => "case", + .k_const => "const", + .k_continue => "continue", + .k_continuing => "continuing", + .k_discard => "discard", + .k_default => "default", + .k_else => "else", + .k_enable => "enable", + .k_f16 => "f16", + .k_f32 => "f32", + .k_fallthrough => "fallthrough", + .k_false => "false", + .k_fn => "fn", + .k_for => "for", + .k_i32 => "i32", + .k_if => "if", + .k_let => "let", + .k_loop => "loop", + .k_mat2x2 => "mat2x2", + .k_mat2x3 => "mat2x3", + .k_mat2x4 => "mat2x4", + .k_mat3x2 => "mat3x2", + .k_mat3x3 => "mat3x3", + .k_mat3x4 => "mat3x4", + .k_mat4x2 => "mat4x2", + .k_mat4x3 => "mat4x3", + .k_mat4x4 => "mat4x4", + .k_override => "override", + .k_ptr => "ptr", + .k_return => "return", + .k_sampler => "sampler", + .k_comparison_sampler => "sampler_comparison", + .k_const_assert => "const_assert", + .k_struct => "struct", + .k_switch => "switch", + .k_texture_depth_2d => "texture_depth_2d", + .k_texture_depth_2d_array => "texture_depth_2d_array", + .k_texture_depth_cube => "texture_depth_cube", + .k_texture_depth_cube_array => "texture_depth_cube_array", + .k_texture_depth_multisampled_2d => "texture_depth_multisampled_2d", + .k_texture_external => "texture_external", + .k_texture_multisampled_2d => "texture_multisampled_2d", + .k_texture_sampled_1d => "texture_1d", + .k_texture_sampled_2d => "texture_2d", + .k_texture_sampled_2d_array => "texture_2d_array", + .k_texture_sampled_3d => "texture_3d", + .k_texture_sampled_cube => "texture_cube", + .k_texture_sampled_cube_array => "texture_cube_array", + .k_texture_storage_1d => "texture_storage_1d", + .k_texture_storage_2d => "texture_storage_2d", + .k_texture_storage_2d_array => "texture_storage_2d_array", + .k_texture_storage_3d => "texture_storage_3d", + .k_true => "true", + .k_type => "type", + .k_u32 => "u32", + .k_var => "var", + .k_vec2 => "vec2", + .k_vec3 => "vec3", + .k_vec4 => "vec4", + .k_while => "while", + }; + } +}; + +pub const keywords = std.ComptimeStringMap(Tag, .{ + .{ "array", .k_array }, + .{ "atomic", .k_atomic }, + .{ "bitcast", .k_bitcast }, + .{ "bool", .k_bool }, + .{ "break", .k_break }, + .{ "case", .k_case }, + .{ "const", .k_const }, + .{ "continue", .k_continue }, + .{ "continuing", .k_continuing }, + .{ "discard", .k_discard }, + .{ "default", .k_default }, + .{ "else", .k_else }, + .{ "enable", .k_enable }, + .{ "f16", .k_f16 }, + .{ "f32", .k_f32 }, + .{ "fallthrough", .k_fallthrough }, + .{ "false", .k_false }, + .{ "fn", .k_fn }, + .{ "for", .k_for }, + .{ "i32", .k_i32 }, + .{ "if", .k_if }, + .{ "let", .k_let }, + .{ "loop", .k_loop }, + .{ "mat2x2", .k_mat2x2 }, + .{ "mat2x3", .k_mat2x3 }, + .{ "mat2x4", .k_mat2x4 }, + .{ "mat3x2", .k_mat3x2 }, + .{ "mat3x3", .k_mat3x3 }, + .{ "mat3x4", .k_mat3x4 }, + .{ "mat4x2", .k_mat4x2 }, + .{ "mat4x3", .k_mat4x3 }, + .{ "mat4x4", .k_mat4x4 }, + .{ "override", .k_override }, + .{ "ptr", .k_ptr }, + .{ "return", .k_return }, + .{ "sampler", .k_sampler }, + .{ "sampler_comparison", .k_comparison_sampler }, + .{ "const_assert", .k_const_assert }, + .{ "struct", .k_struct }, + .{ "switch", .k_switch }, + .{ "texture_depth_2d", .k_texture_depth_2d }, + .{ "texture_depth_2d_array", .k_texture_depth_2d_array }, + .{ "texture_depth_cube", .k_texture_depth_cube }, + .{ "texture_depth_cube_array", .k_texture_depth_cube_array }, + .{ "texture_depth_multisampled_2d", .k_texture_depth_multisampled_2d }, + .{ "texture_external", .k_texture_external }, + .{ "texture_multisampled_2d", .k_texture_multisampled_2d }, + .{ "texture_1d", .k_texture_sampled_1d }, + .{ "texture_2d", .k_texture_sampled_2d }, + .{ "texture_2d_array", .k_texture_sampled_2d_array }, + .{ "texture_3d", .k_texture_sampled_3d }, + .{ "texture_cube", .k_texture_sampled_cube }, + .{ "texture_cube_array", .k_texture_sampled_cube_array }, + .{ "texture_storage_1d", .k_texture_storage_1d }, + .{ "texture_storage_2d", .k_texture_storage_2d }, + .{ "texture_storage_2d_array", .k_texture_storage_2d_array }, + .{ "texture_storage_3d", .k_texture_storage_3d }, + .{ "true", .k_true }, + .{ "type", .k_type }, + .{ "u32", .k_u32 }, + .{ "var", .k_var }, + .{ "vec2", .k_vec2 }, + .{ "vec3", .k_vec3 }, + .{ "vec4", .k_vec4 }, + .{ "while", .k_while }, +}); diff --git a/libs/dusk/src/Tokenizer.zig b/libs/dusk/src/Tokenizer.zig new file mode 100644 index 00000000..8a26decc --- /dev/null +++ b/libs/dusk/src/Tokenizer.zig @@ -0,0 +1,417 @@ +const std = @import("std"); +const Token = @import("Token.zig"); + +const Tokenizer = @This(); + +source: [:0]const u8, +index: u32, + +const State = enum { + start, + invalid, + ident, + underscore, + number, + block_comment, + ampersand, + bang, + equal, + greater, + shift_right, + less, + shift_left, + minus, + mod, + pipe, + plus, + slash, + star, + xor, +}; + +pub fn dump(self: *Tokenizer, token: Token) void { + std.debug.print("\x1b[0;33m{s} \x1b[0;90m\"{s}\"\x1b[0m\n", .{ @tagName(token.tag), token.loc.slice(self.source) }); +} + +pub fn init(source: [:0]const u8) Tokenizer { + // Skip the UTF-8 BOM if present + const src_start: u32 = if (std.mem.startsWith(u8, source, "\xEF\xBB\xBF")) 3 else 0; + return Tokenizer{ + .source = source[src_start..], + .index = 0, + }; +} + +pub fn peek(self: *Tokenizer) Token { + var index = self.index; + var state = State.start; + var result = Token{ + .tag = .eof, + .loc = .{ + .start = index, + .end = undefined, + }, + }; + + while (true) : (index += 1) { + const c = self.source[index]; + switch (state) { + .start => switch (c) { + 0 => { + if (index != self.source.len) { + result.tag = .invalid; + result.loc.start = index; + index += 1; + result.loc.end = index; + return result; + } + break; + }, + ' ', '\n', '\t', '\r' => result.loc.start = index + 1, + 'a'...'z', 'A'...'Z' => state = .ident, + '0'...'9' => state = .number, + + '&' => state = .ampersand, + '!' => state = .bang, + '=' => state = .equal, + '>' => state = .greater, + '<' => state = .less, + '-' => state = .minus, + '%' => state = .mod, + '|' => state = .pipe, + '+' => state = .plus, + '/' => state = .slash, + '*' => state = .star, + '_' => state = .underscore, + '^' => state = .xor, + + '@' => { + result.tag = .attr; + index += 1; + break; + }, + '[' => { + result.tag = .bracket_left; + index += 1; + break; + }, + ']' => { + result.tag = .bracket_right; + index += 1; + break; + }, + '{' => { + result.tag = .brace_left; + index += 1; + break; + }, + '}' => { + result.tag = .brace_right; + index += 1; + break; + }, + ':' => { + result.tag = .colon; + index += 1; + break; + }, + ',' => { + result.tag = .comma; + index += 1; + break; + }, + '(' => { + result.tag = .paren_left; + index += 1; + break; + }, + ')' => { + result.tag = .paren_right; + index += 1; + break; + }, + '.' => { + result.tag = .period; + index += 1; + break; + }, + ';' => { + result.tag = .semicolon; + index += 1; + break; + }, + '~' => { + result.tag = .tilde; + index += 1; + break; + }, + + else => { + state = .invalid; + result.tag = .invalid; + }, + }, + .invalid => break, + + .ident => switch (c) { + 'a'...'z', 'A'...'Z', '0'...'9', '_' => {}, + else => { + result.tag = .ident; + if (Token.keywords.get(self.source[result.loc.start..index])) |tag| { + result.tag = tag; + } + break; + }, + }, + .underscore => switch (c) { // TODO: two underscore `__` https://www.w3.org/TR/WGSL/#identifiers + 'a'...'z', 'A'...'Z', '_', '0'...'9' => state = .ident, + else => { + result.tag = .underscore; + break; + }, + }, + + .number => switch (c) { + '0'...'9', '.', 'i', 'u', 'f', 'h', 'e', '-', '+' => {}, + else => { + result.tag = .number; + break; + }, + }, + + .block_comment => switch (c) { + 0 => break, + '\n' => { + state = .start; + result.loc.start = index + 1; + }, + else => {}, + }, + + .ampersand => switch (c) { + '&' => { + result.tag = .and_and; + index += 1; + break; + }, + '=' => { + result.tag = .and_equal; + index += 1; + break; + }, + else => { + result.tag = .@"and"; + break; + }, + }, + .bang => switch (c) { + '=' => { + result.tag = .not_equal; + index += 1; + break; + }, + else => { + result.tag = .bang; + break; + }, + }, + .equal => switch (c) { + '=' => { + result.tag = .equal_equal; + index += 1; + break; + }, + else => { + result.tag = .equal; + break; + }, + }, + .greater => switch (c) { + '>' => state = .shift_right, + '=' => { + result.tag = .greater_than_equal; + index += 1; + break; + }, + else => { + result.tag = .greater_than; + break; + }, + }, + .shift_right => switch (c) { + '=' => { + result.tag = .shift_right_equal; + index += 1; + break; + }, + else => { + result.tag = .shift_right; + break; + }, + }, + .less => switch (c) { + '<' => state = .shift_left, + '=' => { + result.tag = .less_than_equal; + index += 1; + break; + }, + else => { + result.tag = .less_than; + break; + }, + }, + .shift_left => switch (c) { + '=' => { + result.tag = .shift_left_equal; + index += 1; + break; + }, + else => { + result.tag = .shift_left; + break; + }, + }, + .minus => switch (c) { + '-' => { + result.tag = .minus_minus; + index += 1; + break; + }, + '=' => { + result.tag = .minus_equal; + index += 1; + break; + }, + '>' => { + result.tag = .arrow; + index += 1; + break; + }, + else => { + result.tag = .minus; + break; + }, + }, + .mod => switch (c) { + '=' => { + result.tag = .modulo_equal; + index += 1; + break; + }, + else => { + result.tag = .mod; + break; + }, + }, + .pipe => switch (c) { + '|' => { + result.tag = .or_or; + index += 1; + break; + }, + '=' => { + result.tag = .or_equal; + index += 1; + break; + }, + else => { + result.tag = .@"or"; + break; + }, + }, + .plus => switch (c) { + '+' => { + result.tag = .plus_plus; + index += 1; + break; + }, + '=' => { + result.tag = .plus_equal; + index += 1; + break; + }, + else => { + result.tag = .plus; + break; + }, + }, + .slash => switch (c) { + '/' => state = .block_comment, + '=' => { + result.tag = .division_equal; + index += 1; + break; + }, + else => { + result.tag = .division; + break; + }, + }, + .star => switch (c) { + '=' => { + result.tag = .times_equal; + index += 1; + break; + }, + else => { + result.tag = .star; + break; + }, + }, + .xor => switch (c) { + '=' => { + result.tag = .xor_equal; + index += 1; + break; + }, + else => { + result.tag = .xor; + break; + }, + }, + } + } + + result.loc.end = index; + return result; +} + +pub fn next(self: *Tokenizer) Token { + const tok = self.peek(); + self.index = tok.loc.end; + return tok; +} + +test "tokenize identifier and numbers" { + comptime var str: [:0]const u8 = + \\_ __ _iden iden 100.8i // cc + \\// commnet + \\ + ; + var tokenizer = Tokenizer.init(str); + try std.testing.expect(tokenizer.next().tag == .underscore); + try std.testing.expect(tokenizer.next().tag == .ident); + try std.testing.expect(tokenizer.next().tag == .ident); + try std.testing.expect(tokenizer.next().tag == .ident); + try std.testing.expect(tokenizer.next().tag == .number); + try std.testing.expect(tokenizer.next().tag == .eof); +} + +test "tokenize other" { + comptime var str: [:0]const u8 = ""; + inline for (std.meta.fields(Token.Tag), 0..) |field, i| comptime { + if (i > 3) { + str = str ++ " " ++ (Token.Tag.symbol(@intToEnum(Token.Tag, field.value))); + } + }; + + var tokenizer = Tokenizer.init(str); + + comptime var i = 4; // skip identifiers and nums + inline while (i < std.meta.fields(Token.Tag).len) : (i += 1) { + const tag = @intToEnum(Token.Tag, i); + try std.testing.expect(tokenizer.next().tag == tag); + } + try std.testing.expect(tokenizer.next().tag == .eof); +} diff --git a/libs/dusk/src/main.zig b/libs/dusk/src/main.zig new file mode 100644 index 00000000..f4b16791 --- /dev/null +++ b/libs/dusk/src/main.zig @@ -0,0 +1,61 @@ +const std = @import("std"); + +pub const Ast = @import("Ast.zig"); +pub const Analyse = @import("Analyse.zig"); +pub const Parser = @import("Parser.zig"); +pub const Token = @import("Token.zig"); +pub const Tokenizer = @import("Tokenizer.zig"); + +pub const Extension = enum { + f16, + + pub const Array = std.enums.EnumArray(Extension, bool); +}; + +pub const ErrorMsg = struct { + loc: Token.Loc, + msg: []const u8, + note: ?Note, + + pub const Note = struct { + loc: ?Token.Loc, + msg: []const u8, + + pub fn create( + allocator: std.mem.Allocator, + loc: ?Token.Loc, + comptime format: []const u8, + args: anytype, + ) !Note { + return .{ + .loc = loc, + .msg = try std.fmt.allocPrint(allocator, comptime format, args), + }; + } + + pub fn deinit(note: *Note, allocator: std.mem.Allocator) void { + allocator.free(note.msg); + note.* = undefined; + } + }; + + pub fn create( + allocator: std.mem.Allocator, + loc: Token.Loc, + comptime format: []const u8, + args: anytype, + note: ?Note, + ) !ErrorMsg { + return .{ + .loc = loc, + .msg = try std.fmt.allocPrint(allocator, comptime format, args), + .note = note, + }; + } + + pub fn deinit(err_msg: *ErrorMsg, allocator: std.mem.Allocator) void { + if (err_msg.note) |*note| note.*.deinit(allocator); + allocator.free(err_msg.msg); + err_msg.* = undefined; + } +}; diff --git a/libs/dusk/test/boids.wgsl b/libs/dusk/test/boids.wgsl new file mode 100644 index 00000000..caa67df7 --- /dev/null +++ b/libs/dusk/test/boids.wgsl @@ -0,0 +1,107 @@ +const NUM_PARTICLES: u32 = 1500u; + +struct Particle { + pos : vec2, + vel : vec2, +} + +struct SimParams { + deltaT : f32, + rule1Distance : f32, + rule2Distance : f32, + rule3Distance : f32, + rule1Scale : f32, + rule2Scale : f32, + rule3Scale : f32, +} + +struct Particles { + particles : array +} + +@group(0) @binding(0) var params : SimParams; +@group(0) @binding(1) var particlesSrc : Particles; +@group(0) @binding(2) var particlesDst : Particles; + +// https://github.com/austinEng/Project6-Vulkan-Flocking/blob/master/data/shaders/computeparticles/particle.comp +@compute @workgroup_size(64) +fn main(@builtin(global_invocation_id) global_invocation_id : vec3) { + let index : u32 = global_invocation_id.x; + if index >= NUM_PARTICLES { + return; + } + + var vPos = particlesSrc.particles[index].pos; + var vVel = particlesSrc.particles[index].vel; + + var cMass = vec2(0.0, 0.0); + var cVel = vec2(0.0, 0.0); + var colVel = vec2(0.0, 0.0); + var cMassCount : i32 = 0; + var cVelCount : i32 = 0; + + var pos : vec2; + var vel : vec2; + var i : u32 = 0u; + loop { + if i >= NUM_PARTICLES { + break; + } + if i == index { + continue; + } + + pos = particlesSrc.particles[i].pos; + vel = particlesSrc.particles[i].vel; + + if distance(pos, vPos) < params.rule1Distance { + cMass = cMass + pos; + cMassCount = cMassCount + 1; + } + if distance(pos, vPos) < params.rule2Distance { + colVel = colVel - (pos - vPos); + } + if distance(pos, vPos) < params.rule3Distance { + cVel = cVel + vel; + cVelCount = cVelCount + 1; + } + + continuing { + i = i + 1u; + } + } + if cMassCount > 0 { + cMass = cMass / f32(cMassCount) - vPos; + } + if cVelCount > 0 { + cVel = cVel / f32(cVelCount); + } + + vVel = vVel + (cMass * params.rule1Scale) + + (colVel * params.rule2Scale) + + (cVel * params.rule3Scale); + + // clamp velocity for a more pleasing simulation + vVel = normalize(vVel) * clamp(length(vVel), 0.0, 0.1); + + // kinematic update + vPos = vPos + (vVel * params.deltaT); + + // Wrap around boundary + if vPos.x < -1.0 { + vPos.x = 1.0; + } + if vPos.x > 1.0 { + vPos.x = -1.0; + } + if vPos.y < -1.0 { + vPos.y = 1.0; + } + if vPos.y > 1.0 { + vPos.y = -1.0; + } + + // Write back + particlesDst.particles[index].pos = vPos; + particlesDst.particles[index].vel = vVel; +} diff --git a/libs/dusk/test/gkurve.wgsl b/libs/dusk/test/gkurve.wgsl new file mode 100644 index 00000000..c17c3d5d --- /dev/null +++ b/libs/dusk/test/gkurve.wgsl @@ -0,0 +1,199 @@ +struct FragUniform { + type_: u32, + padding: vec3, + blend_color: vec4, +} +@binding(1) @group(0) var ubos: array; +@binding(2) @group(0) var mySampler: sampler; +@binding(3) @group(0) var myTexture: texture_2d; + +const wireframe = false; +const antialiased = true; +const aa_px = 1.0; // pixels to consume for AA +const dist_scale_px = 300.0; // TODO: do not hard code + +@fragment fn main( + @location(0) uv: vec2, + @interpolate(linear) @location(1) bary_in: vec2, + @interpolate(flat) @location(2) triangle_index: u32, +) -> @location(0) vec4 { + // Example 1: Visualize barycentric coordinates: + // let bary = bary_in; + // return vec4(bary.x, bary.y, 0.0, 1.0); + // return vec4(0.0, bary.x, 0.0, 1.0); // [1.0 (bottom-left vertex), 0.0 (bottom-right vertex)] + // return vec4(0.0, bary.y, 0.0, 1.0); // [1.0 (bottom-left vertex), 0.0 (top-right face)] + + // Example 2: Very simple quadratic bezier + // let bary = bary_in; + // if (bary.x * bary.x - bary.y) > 0 { + // discard; + // } + // return vec4(0.0, 1.0, 0.0, 1.0); + + // Example 3: Render gkurve primitives + let inversion = select( 1.0, -1.0, ubos[triangle_index].type_ == 0u || ubos[triangle_index].type_ == 1u); + // Texture uvs + var correct_uv = uv; + correct_uv.y = 1.0 - correct_uv.y; + var color = textureSample(myTexture, mySampler, correct_uv) * ubos[triangle_index].blend_color; + + // Curve rendering + let border_color = vec4(1.0, 0.0, 0.0, 1.0); + let border_px = 30.0; + let is_semicircle = ubos[triangle_index].type_ == 1u || ubos[triangle_index].type_ == 3u; + var result = select( + curveColor(bary_in, border_px, border_color, color, inversion, is_semicircle), + color, + ubos[triangle_index].type_ == 4u, // triangle rendering + ); + + // Wireframe rendering + let wireframe_px = 1.0; + let wireframe_color = vec4(0.5, 0.5, 0.5, 1.0); + if (wireframe) { + result = wireframeColor(bary_in, wireframe_px, wireframe_color, result); + } + + if (result.a == 0.0) { discard; } + return result; +} + +// Performs alpha 'over' blending between two premultiplied-alpha colors. +fn alphaOver(a: vec4, b: vec4) -> vec4 { + return a + (b * (1.0 - a.a)); +} + +// Calculates signed distance to a quadratic bézier curve using barycentric coordinates. +fn distanceToQuadratic(bary: vec2) -> f32 { + // Gradients + let px = dpdx(bary.xy); + let py = dpdy(bary.xy); + + // Chain rule + let fx = (2.0 * bary.x) * px.x - px.y; + let fy = (2.0 * bary.x) * py.x - py.y; + + return (bary.x * bary.x - bary.y) / sqrt(fx * fx + fy * fy); +} + +// Calculates signed distance to a semicircle using barycentric coordinates. +fn distanceToSemicircle(bary: vec2) -> f32 { + let x = abs(((bary.x - 0.5) * 2.0)); // [0.0 left, 1.0 center, 0.0 right] + let y = ((bary.x-bary.y) * 4.0); // [2.0 bottom, 0.0 top] + let c = x*x + y*y; + + // Gradients + let px = dpdx(bary.xy); + let py = dpdy(bary.xy); + + // Chain rule + let fx = c * px.x - px.y; + let fy = c * py.x - py.y; + + let d = (1.0 - (x*x + y*y)) - 0.2; + return (-d / 6.0) / sqrt(fx * fx + fy * fy); +} + +// Calculates signed distance to the wireframe (i.e. faces) of the triangle using barycentric +// coordinates. +fn distanceToWireframe(bary: vec2) -> f32 { + let normal = vec3( + bary.y, // distance to right face + (bary.x - bary.y) * 2.0, // distance to bottom face + 1.0 - (((bary.x - bary.y)) + bary.x), // distance to left face + ); + let fw = sqrt(dpdx(normal)*dpdx(normal) + dpdy(normal)*dpdy(normal)); + let d = normal / fw; + return min(min(d.x, d.y), d.z); +} + +// Calculates the color of the wireframe, taking into account antialiasing and alpha blending with +// the desired background blend color. +fn wireframeColor(bary: vec2, px: f32, color: vec4, blend_color: vec4) -> vec4 { + let dist = distanceToWireframe(bary); + if (antialiased) { + let outer = dist; + let inner = (px + (aa_px * 2.0)) - dist; + let in_wireframe = outer >= 0.0 && inner >= 0.0; + if (in_wireframe) { + // Note: If this is the outer edge of the wireframe, we do not want to perform alpha + // blending with the background blend color, since it is an antialiased edge and should + // be transparent. However, if it is the internal edge of the wireframe, we do want to + // perform alpha blending as it should be an overlay, not transparent. + let is_outer_edge = outer < inner; + if (is_outer_edge) { + let alpha = smoothstep(0.0, 1.0, outer*(1.0 / aa_px)); + return vec4((color.rgb/color.a)*alpha, alpha); + } else { + let aa_inner = inner - aa_px; + let alpha = smoothstep(0.0, 1.0, aa_inner*(1.0 / aa_px)); + let wireframe_color = vec4((color.rgb/color.a)*alpha, alpha); + return alphaOver(wireframe_color, blend_color); + } + } + return blend_color; + } else { + // If we're at the edge use the wireframe color, otherwise use the background blend_color. + return select(blend_color, color, (px - dist) >= 0.0); + } +} + +// Calculates the color for a curve, taking into account antialiasing and alpha blending with +// the desired background blend color. +// +// inversion: concave (-1.0) or convex (1.0) +// is_semicircle: quadratic bezier (false) or semicircle (true) +fn curveColor( + bary: vec2, + border_px: f32, + border_color: vec4, + blend_color: vec4, + inversion: f32, + is_semicircle: bool, +) -> vec4 { + let dist = select( + distanceToQuadratic(bary), + distanceToSemicircle(bary), + is_semicircle, + ) * inversion; + let is_inverted = (inversion + 1.0) / 2.0; // 1.0 if inverted, 0.0 otherwise + + if (antialiased) { + let outer = dist + ((border_px + (aa_px * 2.0)) * is_inverted); // bottom + let inner = ((border_px + (aa_px * 2.0)) * (1.0-is_inverted)) - dist; // top + let in_border = outer >= 0.0 && inner >= 0.0; + if (in_border) { + // Note: If this is the outer edge of the curve, we do not want to perform alpha + // blending with the background blend color, since it is an antialiased edge and should + // be transparent. However, if it is the internal edge of the curve, we do want to + // perform alpha blending as it should be an overlay, not transparent. + let is_outer_edge = outer < inner; + if (is_outer_edge) { + let aa_outer = outer - (aa_px * is_inverted); + let alpha = smoothstep(0.0, 1.0, aa_outer*(1.0 / aa_px)); + return vec4((border_color.rgb/border_color.a)*alpha, alpha); + } else { + let aa_inner = inner - (aa_px * (1.0 - is_inverted)); + let alpha = smoothstep(0.0, 1.0, aa_inner*(1.0 / aa_px)); + let new_border_color = vec4((border_color.rgb/border_color.a)*alpha, alpha); + return alphaOver(new_border_color, blend_color); + } + return border_color; + } else if (outer >= 0.0) { + return blend_color; + } else { + return vec4(0.0); + } + } else { + let outer = dist + (border_px * is_inverted); + let inner = (border_px * (1.0-is_inverted)) - dist; + let in_border = outer >= 0.0 && inner >= 0.0; + if (in_border) { + return border_color; + } else if (outer >= 0.0) { + return blend_color; + } else { + return vec4(0.0); + } + } +} \ No newline at end of file diff --git a/libs/dusk/test/main.zig b/libs/dusk/test/main.zig new file mode 100644 index 00000000..3eb9b885 --- /dev/null +++ b/libs/dusk/test/main.zig @@ -0,0 +1,170 @@ +const std = @import("std"); +const dusk = @import("dusk"); +const expect = std.testing.expect; +const allocator = std.testing.allocator; + +fn sdkPath(comptime suffix: []const u8) []const u8 { + if (suffix[0] != '/') @compileError("suffix must be an absolute path"); + return comptime blk: { + const root_dir = std.fs.path.dirname(@src().file) orelse "."; + break :blk root_dir ++ suffix; + }; +} + +// TODO: move this to cli/main.zig +pub fn printErrors(errors: []dusk.ErrorMsg, source: []const u8, file_path: ?[]const u8) !void { + var bw = std.io.bufferedWriter(std.io.getStdErr().writer()); + const b = bw.writer(); + const term = std.debug.TTY.Config{ .escape_codes = {} }; + + for (errors) |*err| { + defer err.deinit(allocator); + + const loc_extra = err.loc.extraInfo(source); + + // 'file:line:column error: ' + try term.setColor(b, .Bold); + try b.print("{?s}:{d}:{d} ", .{ file_path, loc_extra.line, loc_extra.col }); + try term.setColor(b, .Red); + try b.writeAll("error: "); + try term.setColor(b, .Reset); + try term.setColor(b, .Bold); + try b.writeAll(err.msg); + try b.writeByte('\n'); + + try printCode(b, term, source, err.loc); + + // note + if (err.note) |note| { + if (note.loc) |note_loc| { + const note_loc_extra = note_loc.extraInfo(source); + + try term.setColor(b, .Reset); + try term.setColor(b, .Bold); + try b.print("{?s}:{d}:{d} ", .{ file_path, note_loc_extra.line, note_loc_extra.col }); + } + try term.setColor(b, .Cyan); + try b.writeAll("note: "); + + try term.setColor(b, .Reset); + try term.setColor(b, .Bold); + try b.writeAll(note.msg); + try b.writeByte('\n'); + + if (note.loc) |note_loc| { + try printCode(b, term, source, note_loc); + } + } + + try term.setColor(b, .Reset); + } + try bw.flush(); +} + +fn printCode(writer: anytype, term: std.debug.TTY.Config, source: []const u8, loc: dusk.Token.Loc) !void { + const loc_extra = loc.extraInfo(source); + try term.setColor(writer, .Dim); + try writer.print("{d} │ ", .{loc_extra.line}); + try term.setColor(writer, .Reset); + try writer.writeAll(source[loc_extra.line_start..loc.start]); + try term.setColor(writer, .Green); + try writer.writeAll(source[loc.start..loc.end]); + try term.setColor(writer, .Reset); + try writer.writeAll(source[loc.end..loc_extra.line_end]); + try writer.writeByte('\n'); + + // location pointer + const line_number_len = (std.math.log10(loc_extra.line) + 1) + 3; + try writer.writeByteNTimes( + ' ', + line_number_len + (loc_extra.col - 1), + ); + try term.setColor(writer, .Bold); + try term.setColor(writer, .Green); + try writer.writeByte('^'); + try writer.writeByteNTimes('~', loc.end - loc.start - 1); + try writer.writeByte('\n'); +} + +fn expectTree(source: [:0]const u8) !dusk.Ast { + var res = try dusk.Ast.parse(allocator, source); + switch (res) { + .tree => |*tree| { + errdefer tree.deinit(allocator); + if (try tree.analyse(allocator)) |errors| { + try printErrors(errors, source, null); + allocator.free(errors); + return error.Analysing; + } + return tree.*; + }, + .errors => |err_msgs| { + try printErrors(err_msgs, source, null); + allocator.free(err_msgs); + return error.Parsing; + }, + } +} + +test "empty" { + const source = ""; + var tree = try expectTree(source); + defer tree.deinit(allocator); +} + +test "boids" { + const source = @embedFile("boids.wgsl"); + var tree = try expectTree(source); + defer tree.deinit(allocator); +} + +test "gkurve" { + if (true) return error.SkipZigTest; + + const source = @embedFile("gkurve.wgsl"); + var tree = try expectTree(source); + defer tree.deinit(allocator); +} + +test "variable & expressions" { + const source = "var expr = 1 + 5 + 2 * 3 > 6 >> 7;"; + + var tree = try expectTree(source); + defer tree.deinit(allocator); + + const root_node = 0; + try expect(tree.nodeLHS(root_node) + 1 == tree.nodeRHS(root_node)); + + const variable = tree.spanToList(root_node)[0]; + const variable_name = tree.tokenLoc(tree.extraData(dusk.Ast.Node.GlobalVarDecl, tree.nodeLHS(variable)).name); + try expect(std.mem.eql(u8, "expr", variable_name.slice(source))); + try expect(tree.nodeTag(variable) == .global_variable); + try expect(tree.tokenTag(tree.nodeToken(variable)) == .k_var); + + const expr = tree.nodeRHS(variable); + try expect(tree.nodeTag(expr) == .greater); + + const @"1 + 5 + 2 * 3" = tree.nodeLHS(expr); + try expect(tree.nodeTag(@"1 + 5 + 2 * 3") == .add); + + const @"1 + 5" = tree.nodeLHS(@"1 + 5 + 2 * 3"); + try expect(tree.nodeTag(@"1 + 5") == .add); + + const @"1" = tree.nodeLHS(@"1 + 5"); + try expect(tree.nodeTag(@"1") == .number_literal); + + const @"5" = tree.nodeRHS(@"1 + 5"); + try expect(tree.nodeTag(@"5") == .number_literal); + + const @"2 * 3" = tree.nodeRHS(@"1 + 5 + 2 * 3"); + try expect(tree.nodeTag(@"2 * 3") == .mul); + + const @"6 >> 7" = tree.nodeRHS(expr); + try expect(tree.nodeTag(@"6 >> 7") == .shift_right); + + const @"6" = tree.nodeLHS(@"6 >> 7"); + try expect(tree.nodeTag(@"6") == .number_literal); + + const @"7" = tree.nodeRHS(@"6 >> 7"); + try expect(tree.nodeTag(@"7") == .number_literal); +}