diff --git a/.github/workflows/build.yml b/.github/workflows/build.yml index b83728ee..e012f81f 100644 --- a/.github/workflows/build.yml +++ b/.github/workflows/build.yml @@ -1,4 +1,5 @@ on: + workflow_dispatch: release: types: [published] push: @@ -35,7 +36,7 @@ jobs: name: Build on ${{ matrix.name }} runs-on: ${{ matrix.os }} steps: - - uses: actions/checkout@v2 + - uses: actions/checkout@v4 with: submodules: true fetch-depth: 1 @@ -90,6 +91,7 @@ jobs: ./LLGI_Test.exe test - name: Upload + if: matrix.name == 'windows_x64' uses: actions/upload-artifact@v4 with: name: Test_Result_Windows diff --git a/.gitignore b/.gitignore index 15a7a582..0e81970f 100644 --- a/.gitignore +++ b/.gitignore @@ -34,6 +34,9 @@ /msvc/*.user **/Debug /build +/build-webgpu-browser/ /build_clangformat +/thirdparty/dawn/ +/node_modules/ .DS_Store diff --git a/CMakeLists.txt b/CMakeLists.txt index 5722d944..edb90f5d 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -5,12 +5,26 @@ project(LLGI) include(GNUInstallDirs) include(ExternalProject) +include(CTest) # linux flag if(UNIX AND NOT APPLE) set(LINUX TRUE) endif() +option(BUILD_WEBGPU "build webgpu backend" OFF) +option(BUILD_WEBGPU_BROWSER_TEST + "build a WebAssembly browser smoke test for the WebGPU backend" OFF) +if(BUILD_WEBGPU_BROWSER_TEST AND NOT BUILD_WEBGPU) + message(FATAL_ERROR "BUILD_WEBGPU_BROWSER_TEST requires BUILD_WEBGPU=ON.") +endif() +set(WEBGPU_DAWN_SOURCE_DIR + "" + CACHE PATH "Path to a Dawn source checkout used when BUILD_WEBGPU is ON") +option(WEBGPU_DAWN_BUILD_SAMPLES "Build Dawn sample executables" OFF) +option(WEBGPU_DAWN_FORCE_SYSTEM_COMPONENT_LOAD + "Let Dawn load required Windows system components from System32" + ON) option(BUILD_VULKAN "build vulkan" OFF) option(BUILD_VULKAN_COMPILER "build vulkan compiler" OFF) option(BUILD_TEST "build test" OFF) @@ -27,7 +41,7 @@ option(SPIRVCROSS_WITHOUT_INSTALL "Compile with spirv-cross without install" OFF) option(USE_CREATE_COMPILER_FUNCTION "Whether LLGI::CreateCompiler is used." ON) -if(LINUX) +if(LINUX AND NOT BUILD_WEBGPU) set(BUILD_VULKAN TRUE) endif() @@ -246,13 +260,46 @@ if(BUILD_VULKAN) endif() endif() +if(BUILD_WEBGPU AND NOT EMSCRIPTEN) + set(DAWN_FETCH_DEPENDENCIES + OFF + CACHE BOOL + "Dawn dependencies should be fetched before configuring LLGI" + FORCE) + set(DAWN_BUILD_SAMPLES ${WEBGPU_DAWN_BUILD_SAMPLES} + CACHE BOOL "Build Dawn samples" FORCE) + set(DAWN_FORCE_SYSTEM_COMPONENT_LOAD + ${WEBGPU_DAWN_FORCE_SYSTEM_COMPONENT_LOAD} + CACHE BOOL "Allow Dawn to load Windows system components from System32" + FORCE) + set(DAWN_BUILD_TESTS OFF CACHE BOOL "Build Dawn tests" FORCE) + set(DAWN_ENABLE_INSTALL OFF CACHE BOOL "Install Dawn targets" FORCE) + + set(LLGI_DAWN_SOURCE_DIR "") + if(WEBGPU_DAWN_SOURCE_DIR) + set(LLGI_DAWN_SOURCE_DIR "${WEBGPU_DAWN_SOURCE_DIR}") + elseif(EXISTS "${CMAKE_CURRENT_SOURCE_DIR}/thirdparty/dawn/CMakeLists.txt") + set(LLGI_DAWN_SOURCE_DIR "${CMAKE_CURRENT_SOURCE_DIR}/thirdparty/dawn") + endif() + + if(LLGI_DAWN_SOURCE_DIR) + add_subdirectory("${LLGI_DAWN_SOURCE_DIR}" "${CMAKE_CURRENT_BINARY_DIR}/dawn" + EXCLUDE_FROM_ALL) + else() + message( + FATAL_ERROR + "Dawn source was not found. Run `python scripts/fetch_dawn.py`, add thirdparty/dawn, or set WEBGPU_DAWN_SOURCE_DIR to a prepared Dawn checkout." + ) + endif() +endif() + if(APPLE) add_compile_definitions(ENABLE_METAL) endif() add_subdirectory("src") -if(BUILD_TEST) +if(BUILD_TEST OR BUILD_WEBGPU_BROWSER_TEST) add_subdirectory("src_test") endif() diff --git a/README.md b/README.md index 9eee953c..ca7116af 100644 --- a/README.md +++ b/README.md @@ -1,48 +1,230 @@ # LLGI -How to build ----------- +LLGI is a cross-platform graphics abstraction library that provides a common +API for DirectX 12, Metal, and Vulkan. -### Windows +This repository currently contains: +- `LLGI`: the core static library +- `LLGI_Test`: rendering and shader tests (`BUILD_TEST=ON`) +- `example_glfw`, `example_imgui`, `example_GPUParticle`: sample applications + (`BUILD_EXAMPLE=ON`) +- `ShaderTranspiler`: shader conversion tool (`BUILD_TOOL=ON`) + +## Backend and platform support + +| Platform | Default backend | Optional backends | Notes | +| --- | --- | --- | --- | +| Windows | DirectX 12 | Vulkan | CI builds both x86 and x64 configurations. | +| macOS | Metal | - | CI builds with `CMAKE_OSX_DEPLOYMENT_TARGET=10.15`. | +| iOS | Metal | - | CI includes an iOS build configuration. | +| Linux | Vulkan | - | `BUILD_VULKAN` is forced to `ON` on Linux. | + +## Repository layout + +| Path | Description | +| --- | --- | +| `src/` | LLGI library sources and public headers | +| `src_test/` | Test executable and shader test assets | +| `examples/` | GLFW, Dear ImGui, and GPU particle samples | +| `tools/` | Shader transpiler sources | +| `scripts/transpile.py` | Helper script to batch-convert shader assets | +| `thirdparty/` | Bundled third-party dependencies used by Vulkan compiler/tool builds | + +## Getting the source + +```bash +git clone https://github.com/effekseer/LLGI.git +cd LLGI +git submodule update --init --recursive ``` -$ git clone https://github.com/altseed/LLGI.git -$ cd LLGI -$ git submodule update --init -$ cmake -S . -B build -DBUILD_TEST=ON -$ cmake --build build + +## Build requirements + +- CMake 3.15 or newer +- A C++ toolchain for your platform (`Visual Studio`, `Xcode`, `clang`, or `gcc`) +- On Linux, Vulkan and X11 development packages +- When `BUILD_VULKAN_COMPILER` or `BUILD_TOOL` is enabled, the bundled + `glslang` and `SPIRV-Cross` submodules are used by default + +Linux packages used in CI: + +```bash +sudo apt-get update +sudo apt-get install -y \ + libx11-dev \ + libxrandr-dev \ + libxi-dev \ + libxinerama-dev \ + libxcursor-dev \ + libudev-dev \ + libx11-xcb-dev \ + libglu1-mesa-dev \ + mesa-common-dev \ + libvulkan-dev ``` -### macOS +## Common CMake options + +| Option | Default | Description | +| --- | --- | --- | +| `BUILD_TEST` | `OFF` | Build `LLGI_Test` | +| `BUILD_EXAMPLE` | `OFF` | Build sample applications | +| `BUILD_TOOL` | `OFF` | Build `ShaderTranspiler` | +| `BUILD_WEBGPU` | `OFF` | Enable the experimental WebGPU backend. See [docs/WebGPU.md](docs/WebGPU.md). | +| `BUILD_VULKAN` | `OFF` (`ON` on Linux) | Enable the Vulkan backend | +| `BUILD_VULKAN_COMPILER` | `OFF` | Enable Vulkan shader compilation support in `LLGI::CreateCompiler` | +| `USE_CREATE_COMPILER_FUNCTION` | `ON` | Keep `LLGI::CreateCompiler` enabled | +| `USE_MSVC_RUNTIME_LIBRARY_DLL` | `ON` | Use the DLL MSVC runtime (`/MD`) | + +## Build +### Windows (DirectX 12) + +```bash +cmake -S . -B build -DBUILD_TEST=ON -DBUILD_EXAMPLE=ON +cmake --build build --config Release ``` -$ git clone https://github.com/altseed/LLGI.git -$ cd LLGI -$ git submodule update --init -$ cmake -S . -B build -G "Xcode" -DBUILD_TEST=ON -$ cmake --build build + +For a 32-bit build: + +```bash +cmake -S . -B build -A Win32 -DBUILD_TEST=ON +cmake --build build --config Release ``` -### Vulkan(Window, Linux) +### Windows (DirectX 12 + Vulkan tools) +```bash +cmake -S . -B build \ + -DBUILD_TEST=ON \ + -DBUILD_EXAMPLE=ON \ + -DBUILD_TOOL=ON \ + -DBUILD_VULKAN=ON \ + -DBUILD_VULKAN_COMPILER=ON +cmake --build build --config Release ``` -$ git clone https://github.com/altseed/LLGI.git -$ cd LLGI -$ git submodule update --init -$ cmake -S . -B build -DBUILD_VULKAN=ON -DBUILD_TEST=ON -$ cmake --build build + +### macOS + +```bash +cmake -S . -B build -G Xcode \ + -DCMAKE_OSX_DEPLOYMENT_TARGET=10.15 \ + -DBUILD_TEST=ON \ + -DBUILD_EXAMPLE=ON +cmake --build build --config Release ``` -Test ----------- +### iOS + +```bash +cmake -S . -B build-ios -G Xcode \ + -DCMAKE_SYSTEM_NAME=iOS \ + -DCMAKE_OSX_DEPLOYMENT_TARGET=14.0 \ + -DCMAKE_OSX_ARCHITECTURES="arm64;x86_64" +cmake --build build-ios --config Release +``` + +### Linux (Vulkan) + +```bash +cmake -S . -B build \ + -DCMAKE_BUILD_TYPE=Release \ + -DBUILD_TEST=ON \ + -DBUILD_EXAMPLE=ON \ + -DBUILD_TOOL=ON \ + -DBUILD_VULKAN_COMPILER=ON +cmake --build build +``` + +On Linux, `BUILD_VULKAN` is enabled automatically by the top-level +`CMakeLists.txt`. + +### WebGPU (experimental) -Run with Vulkan +The WebGPU backend uses Dawn and is still experimental. Build instructions, +Dawn setup, test commands, shader notes, and current limitations are documented +in [docs/WebGPU.md](docs/WebGPU.md). + +## Install + +```bash +cmake --install build --prefix ``` -./LLGI_Test --vulkan + +For multi-config generators such as Visual Studio or Xcode, add +`--config Release`. + +The install step exports the `LLGI` static library, public headers, and CMake +package files under `lib/cmake`. + +## Running tests + +`LLGI_Test` is available when `BUILD_TEST=ON`. + +Default device selection: + +- Windows: DirectX 12 +- macOS/iOS: Metal +- Linux: Vulkan + +Examples: + +```bash +# Linux / macOS +./build/src_test/LLGI_Test +./build/src_test/LLGI_Test --filter=SimpleRender.* + +# Windows +build\src_test\Release\LLGI_Test.exe +build\src_test\Release\LLGI_Test.exe --filter=Compile.* +build\src_test\Release\LLGI_Test.exe --vulkan +build\src_test\Release\LLGI_Test.exe --webgpu ``` -Run with single test +If you want Vulkan shader compilation through +`LLGI::CreateCompiler(DeviceType::Vulkan)` or the `Compile.*` tests on Vulkan, +configure with `-DBUILD_VULKAN_COMPILER=ON`. + +## Examples + +When `BUILD_EXAMPLE=ON`, the following targets are built: +- `example_glfw`: minimal clear/present flow using GLFW +- `example_imgui`: Dear ImGui integration on top of LLGI and GLFW +- `example_GPUParticle`: GPU particle sample + +The smallest end-to-end sample is in `examples/glfw/main.cpp`. + +## Shader tools + +When `BUILD_TOOL=ON`, the repository builds `ShaderTranspiler`. + +The helper script below uses the built tool to batch-convert shader assets: + +```bash +python scripts/transpile.py src_test/Shaders/ +python scripts/transpile.py examples/GPUParticle/Shaders/ ``` -./LLGI_Test --filter= + +Additional notes are available in [tools/README.md](tools/README.md). + +## Minimal usage + +```cpp +auto window = std::unique_ptr(LLGI::CreateWindow("LLGI", {1280, 720})); + +LLGI::PlatformParameter parameter; +parameter.Device = LLGI::DeviceType::Default; + +auto platform = LLGI::CreateSharedPtr(LLGI::CreatePlatform(parameter, window.get())); +auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); +auto memoryPool = LLGI::CreateSharedPtr(graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128)); +auto commandList = LLGI::CreateSharedPtr(graphics->CreateCommandList(memoryPool.get())); ``` + +For a complete runnable example, see `examples/glfw/main.cpp`. + +## License + +LLGI is distributed under the zlib license. See [LICENSE](LICENSE). diff --git a/docs/WebGPU.md b/docs/WebGPU.md new file mode 100644 index 00000000..33180cad --- /dev/null +++ b/docs/WebGPU.md @@ -0,0 +1,153 @@ +# WebGPU Backend + +The WebGPU backend is experimental. It uses Dawn as the native WebGPU +implementation and WGSL as the shader source format. + +## Requirements + +- CMake 3.15 or newer +- A C++17-capable toolchain +- `BUILD_WEBGPU=ON` +- Dawn, prepared with `scripts/fetch_dawn.py` or supplied as an existing checkout +- `BUILD_TOOL=ON` when rebuilding WGSL shader assets with `ShaderTranspiler` + +On Windows, Dawn normally uses D3D12. Vulkan or other Dawn adapters may be +available depending on the local Dawn build and runtime environment. + +## CMake Options + +| Option | Default | Description | +| --- | --- | --- | +| `BUILD_WEBGPU` | `OFF` | Enable the WebGPU backend | +| `WEBGPU_DAWN_SOURCE_DIR` | empty | Use an existing Dawn checkout | +| `WEBGPU_DAWN_BUILD_SAMPLES` | `OFF` | Build Dawn sample executables | +| `WEBGPU_DAWN_FORCE_SYSTEM_COMPONENT_LOAD` | `ON` | Let Dawn load Windows system components such as `d3dcompiler_47.dll` from System32 | + +## Getting Dawn + +### Recommended Flow + +Dawn should be fetched before configuring LLGI. The helper script clones Dawn +into `thirdparty/dawn` and runs Dawn's dependency fetch script: + +```bash +python scripts/fetch_dawn.py +``` + +Pin a Dawn revision for reproducible builds: + +```bash +python scripts/fetch_dawn.py --revision +``` + +Then configure LLGI: + +```bash +cmake -S . -B build-webgpu \ + -DBUILD_WEBGPU=ON \ + -DBUILD_TEST=ON \ + -DBUILD_TOOL=ON +cmake --build build-webgpu --config Release +``` + +### Existing Dawn Checkout + +You can use a separate Dawn checkout if its dependencies have already been +fetched: + +```bash +cmake -S . -B build-webgpu \ + -DBUILD_WEBGPU=ON \ + -DBUILD_TEST=ON \ + -DBUILD_TOOL=ON \ + -DWEBGPU_DAWN_SOURCE_DIR=/path/to/dawn +cmake --build build-webgpu --config Release +``` + +### `thirdparty/dawn` + +The repository checks `thirdparty/dawn` automatically. This is useful when you +want Dawn to live inside the LLGI working tree. Prefer the helper script above, +or run Dawn's official setup manually: + +```bash +cd thirdparty +git clone https://dawn.googlesource.com/dawn dawn +cd dawn +cp scripts/standalone.gclient .gclient +gclient sync +``` + +Then configure LLGI with `-DBUILD_WEBGPU=ON`. + +## Running Tests + +Build `LLGI_Test` with `BUILD_TEST=ON`, then pass `--webgpu`: + +```bash +# Windows +build-webgpu\src_test\Release\LLGI_Test.exe --webgpu +build-webgpu\src_test\Release\LLGI_Test.exe --webgpu --filter=SimpleRender.* + +# Linux / macOS +./build-webgpu/src_test/LLGI_Test --webgpu +./build-webgpu/src_test/LLGI_Test --webgpu --filter=SimpleRender.* +``` + +Some environments need a visible GPU session for Dawn to create a WebGPU device. +Headless CI may need Dawn-specific setup. + +## Browser Smoke Test + +The native test path above uses Dawn directly. To compile the WebGPU backend to +WebAssembly and run it in a real browser WebGPU implementation, use the browser +test flow in `docs/WebGPU_Browser_Test.md`. + +## Shader Generation + +`ShaderTranspiler` can emit WGSL and compiled WGSL blobs for WebGPU tests. + +```bash +cmake -S . -B build-webgpu \ + -DBUILD_TOOL=ON \ + -DBUILD_WEBGPU=ON +cmake --build build-webgpu --config Release --target ShaderTranspiler + +python scripts/transpile.py src_test/Shaders/ +``` + +Generated WebGPU shaders are stored under: + +- `src_test/Shaders/WebGPU/` +- `src_test/Shaders/WebGPU_Compiled/` + +The compiled files use an LLGI header followed by WGSL text. Runtime WebGPU +shader creation expects WGSL or this compiled WGSL format; legacy runtime WGSL +rewrites are not applied. + +## Current Limitations + +- The backend is experimental and is not enabled by default. +- Dawn API and Tint WGSL output can change. Prefer pinning + `scripts/fetch_dawn.py --revision` for stable builds. +- The WebGPU backend currently depends on Dawn CMake targets + `dawn::webgpu_dawn` or `webgpu_dawn`. +- WebGPU shader assets should be regenerated when the shader transpiler or Tint + revision changes. +- `LLGI::CreateCompiler(DeviceType::WebGPU)` is not a runtime HLSL compiler; use + `ShaderTranspiler` to generate WGSL assets. +- External-device construction APIs can run without an owned `wgpu::Instance`. + In that mode, some wait processing is limited to the supplied Dawn objects. + +## Troubleshooting + +- If configure fails because Dawn is missing, run `python scripts/fetch_dawn.py`, + add `thirdparty/dawn`, or set `WEBGPU_DAWN_SOURCE_DIR`. +- If Dawn dependency sync fails, install `depot_tools` and ensure `gclient` is + available on `PATH`, or use a Dawn checkout whose dependencies are already + synced. +- If WebGPU device creation fails on Windows with `d3dcompiler_47.dll`, rebuild + with `WEBGPU_DAWN_FORCE_SYSTEM_COMPONENT_LOAD=ON`. +- If shader creation fails, regenerate WGSL with the current `ShaderTranspiler` + and check that the generated shaders are under `src_test/Shaders/WebGPU/` and + `src_test/Shaders/WebGPU_Compiled/`. diff --git a/docs/WebGPU_Browser_Test.md b/docs/WebGPU_Browser_Test.md new file mode 100644 index 00000000..00417ef3 --- /dev/null +++ b/docs/WebGPU_Browser_Test.md @@ -0,0 +1,320 @@ +# WebGPU Browser Test + +This document describes how to build and run the browser-only WebGPU backend +tests. These tests compile LLGI to WebAssembly with Emscripten, run it in a real +Chromium-family browser, and use Emdawnwebgpu to access the browser WebGPU API. + +The native WebGPU path in `docs/WebGPU.md` uses Dawn directly. The browser path +covered here verifies a separate runtime: + +- Emscripten compilation and linking +- Emdawnwebgpu C/C++ bindings +- browser `navigator.gpu` adapter/device creation +- WGSL shader module creation +- render and compute pipeline creation +- browser canvas surface presentation +- offscreen render target readback +- storage buffer compute readback + +## Requirements + +- CMake 3.15 or newer +- Git +- Node.js +- Emscripten 4.x or newer with `--use-port=emdawnwebgpu` +- Playwright +- A WebGPU-capable Chromium, Chrome, or Edge browser + +WebGPU requires a secure context. The runner serves the generated files from +`localhost`; do not open `LLGI_Test.html` directly with `file://`. + +## Install Emscripten + +Install emsdk from the official repository: + +```bash +git clone https://github.com/emscripten-core/emsdk.git +cd emsdk +./emsdk install latest +./emsdk activate latest +source ./emsdk_env.sh +``` + +On Windows PowerShell: + +```powershell +git clone https://github.com/emscripten-core/emsdk.git +cd emsdk +.\emsdk install latest +.\emsdk activate latest +.\emsdk_env.ps1 +``` + +Check the installation: + +```bash +emcc --version +emcmake --version +``` + +Run the emsdk environment script again whenever you open a new shell. + +## Install Playwright + +From the LLGI repository root: + +```bash +npm install playwright +npx playwright install chromium +``` + +On Linux CI or minimal Linux installations, install browser system dependencies: + +```bash +npx playwright install-deps chromium +``` + +## Configure + +Use Emscripten's CMake wrapper: + +```bash +emcmake cmake -S . -B build-webgpu-browser \ + -DBUILD_WEBGPU=ON \ + -DBUILD_WEBGPU_BROWSER_TEST=ON \ + -DBUILD_TEST=ON +``` + +PowerShell equivalent: + +```powershell +emcmake cmake -S . -B build-webgpu-browser ` + -DBUILD_WEBGPU=ON ` + -DBUILD_WEBGPU_BROWSER_TEST=ON ` + -DBUILD_TEST=ON +``` + +`BUILD_WEBGPU_BROWSER_TEST=ON` requires an Emscripten toolchain and builds a +browser-focused `LLGI_Test.html`. It does not require a native Dawn checkout. + +## Build + +```bash +cmake --build build-webgpu-browser --target LLGI_Test +``` + +The generated files are: + +```text +build-webgpu-browser/src_test/LLGI_Test.html +build-webgpu-browser/src_test/LLGI_Test.js +build-webgpu-browser/src_test/LLGI_Test.wasm +build-webgpu-browser/src_test/LLGI_Test.data +``` + +On Windows, if emsdk is installed on a non-`C:` drive, keep `EM_CACHE` on the +same drive as emsdk. This avoids Emscripten port-build failures caused by +cross-drive relative paths: + +```powershell +$env:EM_CACHE = "D:\emscripten-cache-llgi" +cmake --build build-webgpu-browser --target LLGI_Test +``` + +For CI, cache `EM_CACHE` between runs to avoid rebuilding Emscripten system +libraries and the Emdawnwebgpu port every time. + +## Run Automated Tests + +Run the Playwright harness from the repository root: + +```bash +node src_test/browser/run_webgpu_browser_test.mjs \ + build-webgpu-browser/src_test/LLGI_Test.html +``` + +PowerShell: + +```powershell +node src_test/browser/run_webgpu_browser_test.mjs ` + build-webgpu-browser/src_test/LLGI_Test.html +``` + +The runner: + +- starts a temporary localhost HTTP server +- disables caching for generated `.html`, `.js`, `.wasm`, and `.data` files +- launches Chromium with WebGPU-friendly flags +- waits for the Emscripten module to report completion +- exits non-zero on failure + +The success log ends with: + +```text +LLGI_TEST_PASS completed +``` + +The default filter is: + +```text +WebGPUBrowser.* +``` + +Run one test with `--filter`: + +```bash +node src_test/browser/run_webgpu_browser_test.mjs \ + build-webgpu-browser/src_test/LLGI_Test.html \ + --filter=WebGPUBrowser.ScreenPresentation +``` + +## View in a Browser + +To see the canvas presentation test, serve the generated files: + +```bash +cd build-webgpu-browser/src_test +python -m http.server 8000 +``` + +Open: + +```text +http://localhost:8000/LLGI_Test.html?filter=WebGPUBrowser.ScreenPresentation +``` + +You should see a blue canvas with a colored polygon. Open browser DevTools and +check the Console for: + +```text +Start : WebGPUBrowser.ScreenPresentation +LLGI_TEST_PASS completed +``` + +## CTest + +When Node.js is found during CMake configure, CMake registers: + +```bash +ctest --test-dir build-webgpu-browser -R LLGI_WebGPU_Browser --output-on-failure +``` + +The CTest entry expects the `playwright` package to be available to Node.js. + +## Current Test Cases + +- `WebGPUBrowser.ComputeCompile` + - loads a WGSL compute shader + - compiles a compute pipeline +- `WebGPUBrowser.ComputeDispatch` + - uploads structured input data + - dispatches a storage-buffer compute shader + - copies output to a readback buffer + - maps and verifies computed values +- `WebGPUBrowser.OffscreenRender` + - creates an offscreen render texture + - loads WGSL vertex/fragment shaders + - compiles a render pipeline + - draws a rectangle +- `WebGPUBrowser.RenderReadback` + - clears an offscreen render target + - copies it to a readback buffer + - verifies pixel values +- `WebGPUBrowser.TextureAndConstantRender` + - uploads texture data with `Queue::WriteTexture` + - renders with texture and sampler bind groups + - renders again with vertex/pixel uniform buffers + - reads back pixels and verifies clear color and rendered output +- `WebGPUBrowser.ScreenPresentation` + - creates a browser canvas WebGPU surface + - renders through `PlatformWebGPU::GetCurrentScreen` + - leaves a visible blue canvas with a colored polygon + +Browser readback uses Asyncify so C++ test code can wait for `MapAsync` and +`Queue::OnSubmittedWorkDone` callbacks while the browser event loop continues to +run. + +## CI Notes + +A typical CI flow is: + +```bash +npm install playwright +npx playwright install chromium +emcmake cmake -S . -B build-webgpu-browser \ + -DBUILD_WEBGPU=ON \ + -DBUILD_WEBGPU_BROWSER_TEST=ON \ + -DBUILD_TEST=ON +cmake --build build-webgpu-browser --target LLGI_Test +node src_test/browser/run_webgpu_browser_test.mjs \ + build-webgpu-browser/src_test/LLGI_Test.html +``` + +The runner passes these Chromium flags: + +- `--enable-unsafe-webgpu` +- `--ignore-gpu-blocklist` +- `--enable-features=Vulkan,UseSkiaRenderer` +- `--use-vulkan=swiftshader` + +These help in headless or GPU-limited environments, but browser policy, +drivers, or CI sandboxing can still disable WebGPU. + +## Troubleshooting + +### `navigator.gpu is not available` + +Use `http://localhost` or HTTPS. WebGPU is unavailable from `file://`. + +Also check that the browser supports WebGPU and is not blocked by local GPU +policy. + +### Playwright fails to launch Chromium + +Install the browser: + +```bash +npx playwright install chromium +``` + +If Playwright's downloaded Chromium cannot launch, use an installed browser: + +```powershell +$env:CHROME_PATH = "C:\Program Files\Google\Chrome\Application\chrome.exe" +node src_test/browser/run_webgpu_browser_test.mjs ` + build-webgpu-browser/src_test/LLGI_Test.html +``` + +Common Windows alternatives: + +```powershell +$env:CHROME_PATH = "C:\Program Files (x86)\Google\Chrome\Application\chrome.exe" +$env:CHROME_PATH = "C:\Program Files\Microsoft\Edge\Application\msedge.exe" +$env:CHROME_PATH = "C:\Program Files (x86)\Microsoft\Edge\Application\msedge.exe" +``` + +### Emdawnwebgpu port build fails with a cross-drive path error + +Put `EM_CACHE` on the same drive as emsdk: + +```powershell +$env:EM_CACHE = "D:\emscripten-cache-llgi" +cmake --build build-webgpu-browser --target LLGI_Test +``` + +### Browser shows a dark or blank page + +Most tests render offscreen and validate results through readback. Only +`WebGPUBrowser.ScreenPresentation` intentionally draws to the visible canvas. + +Open: + +```text +http://localhost:8000/LLGI_Test.html?filter=WebGPUBrowser.ScreenPresentation +``` + +### Logs show `LLGI_TEST_FAIL` + +Check the preceding browser console lines. The test code prints `Abort on + : ` for assertion failures, and the runner also reports WebGPU +validation errors captured by the browser. diff --git a/examples/CMakeLists.txt b/examples/CMakeLists.txt index 2516de19..1d735f8a 100644 --- a/examples/CMakeLists.txt +++ b/examples/CMakeLists.txt @@ -1,3 +1,8 @@ +if(UNIX AND NOT APPLE) + set(GLFW_BUILD_WAYLAND OFF CACHE BOOL "Build support for Wayland" FORCE) + set(GLFW_BUILD_X11 ON CACHE BOOL "Build support for X11" FORCE) +endif() + add_subdirectory("thirdparty/glfw/") add_subdirectory("thirdparty/imgui/") add_subdirectory("ImGuiPlatform") diff --git a/examples/GPUParticle/GPUParticle.cpp b/examples/GPUParticle/GPUParticle.cpp index 1064c554..102404b1 100644 --- a/examples/GPUParticle/GPUParticle.cpp +++ b/examples/GPUParticle/GPUParticle.cpp @@ -200,12 +200,31 @@ std::vector Shader::LoadData(const char* path) return ret; } - fseek(fp, 0, SEEK_END); - auto size = ftell(fp); - fseek(fp, 0, SEEK_SET); + if (fseek(fp, 0, SEEK_END) != 0) + { + fclose(fp); + return ret; + } + + const auto size = ftell(fp); + if (size <= 0) + { + fclose(fp); + return ret; + } - ret.resize(size); - fread(ret.data(), 1, size, fp); + if (fseek(fp, 0, SEEK_SET) != 0) + { + fclose(fp); + return ret; + } + + ret.resize(static_cast(size)); + const auto readSize = fread(ret.data(), 1, ret.size(), fp); + if (readSize != ret.size()) + { + ret.resize(readSize); + } fclose(fp); return ret; diff --git a/examples/thirdparty/glfw b/examples/thirdparty/glfw index 7b6aead9..b00e6a8a 160000 --- a/examples/thirdparty/glfw +++ b/examples/thirdparty/glfw @@ -1 +1 @@ -Subproject commit 7b6aead9fb88b3623e3b3725ebb42670cbe4c579 +Subproject commit b00e6a8a88ad1b60c0a045e696301deb92c9a13e diff --git a/package-lock.json b/package-lock.json new file mode 100644 index 00000000..f1e3a8fc --- /dev/null +++ b/package-lock.json @@ -0,0 +1,56 @@ +{ + "name": "LLGI", + "lockfileVersion": 3, + "requires": true, + "packages": { + "": { + "dependencies": { + "playwright": "^1.59.1" + } + }, + "node_modules/fsevents": { + "version": "2.3.2", + "resolved": "https://registry.npmjs.org/fsevents/-/fsevents-2.3.2.tgz", + "integrity": "sha512-xiqMQR4xAeHTuB9uWm+fFRcIOgKBMiOBP+eXiyT7jsgVCq1bkVygt00oASowB7EdtpOHaaPgKt812P9ab+DDKA==", + "hasInstallScript": true, + "license": "MIT", + "optional": true, + "os": [ + "darwin" + ], + "engines": { + "node": "^8.16.0 || ^10.6.0 || >=11.0.0" + } + }, + "node_modules/playwright": { + "version": "1.59.1", + "resolved": "https://registry.npmjs.org/playwright/-/playwright-1.59.1.tgz", + "integrity": "sha512-C8oWjPR3F81yljW9o5OxcWzfh6avkVwDD2VYdwIGqTkl+OGFISgypqzfu7dOe4QNLL2aqcWBmI3PMtLIK233lw==", + "license": "Apache-2.0", + "dependencies": { + "playwright-core": "1.59.1" + }, + "bin": { + "playwright": "cli.js" + }, + "engines": { + "node": ">=18" + }, + "optionalDependencies": { + "fsevents": "2.3.2" + } + }, + "node_modules/playwright-core": { + "version": "1.59.1", + "resolved": "https://registry.npmjs.org/playwright-core/-/playwright-core-1.59.1.tgz", + "integrity": "sha512-HBV/RJg81z5BiiZ9yPzIiClYV/QMsDCKUyogwH9p3MCP6IYjUFu/MActgYAvK0oWyV9NlwM3GLBjADyWgydVyg==", + "license": "Apache-2.0", + "bin": { + "playwright-core": "cli.js" + }, + "engines": { + "node": ">=18" + } + } + } +} diff --git a/package.json b/package.json new file mode 100644 index 00000000..8a8fbea3 --- /dev/null +++ b/package.json @@ -0,0 +1,5 @@ +{ + "dependencies": { + "playwright": "^1.59.1" + } +} diff --git a/scripts/fetch_dawn.py b/scripts/fetch_dawn.py new file mode 100644 index 00000000..203953cd --- /dev/null +++ b/scripts/fetch_dawn.py @@ -0,0 +1,80 @@ +#!/usr/bin/env python3 + +import argparse +import os +import subprocess +import sys + + +def run(args, cwd=None): + print("+ " + " ".join(args)) + subprocess.check_call(args, cwd=cwd) + + +def is_dawn_checkout(path): + return os.path.isfile(os.path.join(path, "CMakeLists.txt")) and os.path.isdir( + os.path.join(path, "src", "dawn") + ) + + +def main(): + parser = argparse.ArgumentParser( + description="Fetch Dawn for the LLGI WebGPU backend." + ) + parser.add_argument( + "-d", + "--directory", + default=os.path.join("thirdparty", "dawn"), + help="Dawn checkout directory. Default: thirdparty/dawn", + ) + parser.add_argument( + "--repository", + default="https://dawn.googlesource.com/dawn", + help="Dawn git repository URL.", + ) + parser.add_argument( + "--revision", + default="main", + help="Dawn branch, tag, or commit to checkout. Default: main", + ) + parser.add_argument( + "--skip-dependencies", + action="store_true", + help="Only clone/update Dawn; do not run Dawn dependency fetch.", + ) + args = parser.parse_args() + + dawn_dir = os.path.abspath(args.directory) + + if os.path.exists(dawn_dir): + if not is_dawn_checkout(dawn_dir): + print( + f"error: {dawn_dir} exists but does not look like a Dawn checkout", + file=sys.stderr, + ) + return 1 + run(["git", "fetch", "--tags", "origin"], cwd=dawn_dir) + else: + parent = os.path.dirname(dawn_dir) + if parent: + os.makedirs(parent, exist_ok=True) + run(["git", "clone", args.repository, dawn_dir]) + + run(["git", "checkout", args.revision], cwd=dawn_dir) + + if not args.skip_dependencies: + dependency_script = os.path.join(dawn_dir, "tools", "fetch_dawn_dependencies.py") + if not os.path.isfile(dependency_script): + print( + f"error: dependency script was not found: {dependency_script}", + file=sys.stderr, + ) + return 1 + run([sys.executable, dependency_script], cwd=dawn_dir) + + print(f"Dawn is ready: {dawn_dir}") + return 0 + + +if __name__ == "__main__": + raise SystemExit(main()) diff --git a/scripts/transpile.py b/scripts/transpile.py index 38ea685e..0b2dbf9c 100644 --- a/scripts/transpile.py +++ b/scripts/transpile.py @@ -23,9 +23,7 @@ elif os.path.isfile(transpiler_path_make): shutil.copy(transpiler_path_make, "./") -transpiler_call = 'ShaderTranspiler' -if platform.system() == 'Linux': - transpiler_call = './ShaderTranspiler' +transpiler_call = os.path.join('.', transpiler_filename) verts = glob.glob(os.path.join(target_directory, 'HLSL_DX12/*.vert'), recursive=True) frags = glob.glob(os.path.join(target_directory, 'HLSL_DX12/*.frag'), recursive=True) @@ -44,6 +42,25 @@ os.makedirs(os.path.join(target_directory, directory), exist_ok=True) subprocess.call([transpiler_call, kind, target, '--input', f, '--output', os.path.join(target_directory, directory, os.path.basename(f))] + ext) +for kind,paths in [ + ('--vert', verts), + ('--frag', frags), + ('--comp', comps) ]: + for f in paths: + os.makedirs(os.path.join(target_directory, 'WebGPU'), exist_ok=True) + os.makedirs(os.path.join(target_directory, 'WebGPU_Compiled'), exist_ok=True) + subprocess.call([ + transpiler_call, + kind, + '-W', + '--input', + f, + '--output', + os.path.join(target_directory, 'WebGPU', os.path.basename(f)), + '--compiled-output', + os.path.join(target_directory, 'WebGPU_Compiled', os.path.basename(f)) + ]) + verts = glob.glob(os.path.join(target_directory, 'GLSL_VULKAN/*.vert'), recursive=True) frags = glob.glob(os.path.join(target_directory, 'GLSL_VULKAN/*.frag'), recursive=True) comps = glob.glob(os.path.join(target_directory, 'GLSL_VULKAN/*.comp'), recursive=True) diff --git a/src/CMakeLists.txt b/src/CMakeLists.txt index 3c9d8305..84703c13 100644 --- a/src/CMakeLists.txt +++ b/src/CMakeLists.txt @@ -12,12 +12,13 @@ if(WIN32) elseif(APPLE) file(GLOB files_mac Mac/*.h Mac/*.cpp Mac/*.mm) list(APPEND files ${files_mac}) +elseif(EMSCRIPTEN) else() file(GLOB files_linux Linux/*.h Linux/*.cpp) list(APPEND files ${files_linux}) endif() -if(MSVC) +if(MSVC AND NOT BUILD_WEBGPU) file(GLOB files_dx12 DX12/*.h DX12/*.cpp) list(APPEND files ${files_dx12}) endif() @@ -28,6 +29,12 @@ if(BUILD_VULKAN) add_definitions(-DENABLE_VULKAN) endif() +if(BUILD_WEBGPU) + file(GLOB files_webgpu WebGPU/*.h WebGPU/*.cpp) + list(APPEND files ${files_webgpu}) + add_definitions(-DENABLE_WEBGPU) +endif() + if(APPLE) file(GLOB files_metal Metal/*.h Metal/*.cpp Metal/*.mm) list(APPEND files ${files_metal}) @@ -67,10 +74,12 @@ add_library(LLGI STATIC ${files}) file(GLOB LOCAL_HEADERS *.h) set_target_properties(LLGI PROPERTIES PUBLIC_HEADER "${LOCAL_HEADERS}") -if(BUILD_VULKAN) -target_compile_features(LLGI PUBLIC cxx_std_17) +if(BUILD_WEBGPU) + target_compile_features(LLGI PUBLIC cxx_std_20) +elseif(BUILD_VULKAN) + target_compile_features(LLGI PUBLIC cxx_std_17) else() -target_compile_features(LLGI PUBLIC cxx_std_14) + target_compile_features(LLGI PUBLIC cxx_std_14) endif() if(BUILD_VULKAN) @@ -84,6 +93,19 @@ if(BUILD_VULKAN) endif() endif() +if(BUILD_WEBGPU) + if(EMSCRIPTEN) + target_compile_options(LLGI PUBLIC --use-port=emdawnwebgpu) + target_link_options(LLGI PUBLIC --use-port=emdawnwebgpu) + elseif(TARGET dawn::webgpu_dawn) + target_link_libraries(LLGI PRIVATE dawn::webgpu_dawn) + elseif(TARGET webgpu_dawn) + target_link_libraries(LLGI PRIVATE webgpu_dawn) + else() + message(FATAL_ERROR "Dawn targets were not found. Expected dawn::webgpu_dawn or webgpu_dawn.") + endif() +endif() + if(WIN32) elseif(APPLE) @@ -101,19 +123,22 @@ endif() # -------------------- # Install -install( - TARGETS LLGI - EXPORT LLGI-export - INCLUDES - DESTINATION include/LLGI - PUBLIC_HEADER DESTINATION include/LLGI - ARCHIVE DESTINATION lib - LIBRARY DESTINATION lib) - -install( - EXPORT LLGI-export - FILE LLGI-config.cmake - DESTINATION lib/cmake - EXPORT_LINK_INTERFACE_LIBRARIES) +# Because dawn doesn't specify export +if(NOT BUILD_WEBGPU) + install( + TARGETS LLGI + EXPORT LLGI-export + INCLUDES + DESTINATION include/LLGI + PUBLIC_HEADER DESTINATION include/LLGI + ARCHIVE DESTINATION lib + LIBRARY DESTINATION lib) + + install( + EXPORT LLGI-export + FILE LLGI-config.cmake + DESTINATION lib/cmake + EXPORT_LINK_INTERFACE_LIBRARIES) +endif() clang_format(LLGI) diff --git a/src/DX12/LLGI.BaseDX12.cpp b/src/DX12/LLGI.BaseDX12.cpp index 8e5e856e..d80e1ff6 100644 --- a/src/DX12/LLGI.BaseDX12.cpp +++ b/src/DX12/LLGI.BaseDX12.cpp @@ -179,6 +179,9 @@ DXGI_FORMAT ConvertFormat(TextureFormatType format) if (format == TextureFormatType::BC3) return DXGI_FORMAT_BC3_UNORM; + if (format == TextureFormatType::BC7) + return DXGI_FORMAT_BC7_UNORM; + if (format == TextureFormatType::BC1_SRGB) return DXGI_FORMAT_BC1_UNORM_SRGB; @@ -188,6 +191,9 @@ DXGI_FORMAT ConvertFormat(TextureFormatType format) if (format == TextureFormatType::BC3_SRGB) return DXGI_FORMAT_BC3_UNORM_SRGB; + if (format == TextureFormatType::BC7_SRGB) + return DXGI_FORMAT_BC7_UNORM_SRGB; + if (format == TextureFormatType::D32) return DXGI_FORMAT_D32_FLOAT; @@ -238,6 +244,9 @@ TextureFormatType ConvertFormat(DXGI_FORMAT format) if (format == DXGI_FORMAT_BC3_UNORM) return TextureFormatType::BC3; + if (format == DXGI_FORMAT_BC7_UNORM) + return TextureFormatType::BC7; + if (format == DXGI_FORMAT_BC1_UNORM_SRGB) return TextureFormatType::BC1_SRGB; @@ -247,6 +256,9 @@ TextureFormatType ConvertFormat(DXGI_FORMAT format) if (format == DXGI_FORMAT_BC3_UNORM_SRGB) return TextureFormatType::BC3_SRGB; + if (format == DXGI_FORMAT_BC7_UNORM_SRGB) + return TextureFormatType::BC7_SRGB; + if (format == DXGI_FORMAT_D32_FLOAT) return TextureFormatType::D32; diff --git a/src/DX12/LLGI.CommandListDX12.cpp b/src/DX12/LLGI.CommandListDX12.cpp index 0bb6a187..f5243823 100644 --- a/src/DX12/LLGI.CommandListDX12.cpp +++ b/src/DX12/LLGI.CommandListDX12.cpp @@ -1010,8 +1010,11 @@ void CommandListDX12::ClearDepth() return; } + auto depthTexture = static_cast(rt->GetDepthTexture()); + const auto clearFlags = HasStencil(depthTexture->GetFormat()) ? D3D12_CLEAR_FLAG_DEPTH | D3D12_CLEAR_FLAG_STENCIL : D3D12_CLEAR_FLAG_DEPTH; + auto handle = rt->GetHandleDSV(); - currentCommandList_->ClearDepthStencilView(handle[0], D3D12_CLEAR_FLAG_DEPTH | D3D12_CLEAR_FLAG_STENCIL, 1.0f, 0, 0, nullptr); + currentCommandList_->ClearDepthStencilView(handle[0], clearFlags, 1.0f, 0, 0, nullptr); } ID3D12GraphicsCommandList* CommandListDX12::GetCommandList() const { return commandList_.get(); } diff --git a/src/DX12/LLGI.GraphicsDX12.cpp b/src/DX12/LLGI.GraphicsDX12.cpp index d26e98c3..3a9a0fd0 100644 --- a/src/DX12/LLGI.GraphicsDX12.cpp +++ b/src/DX12/LLGI.GraphicsDX12.cpp @@ -4,10 +4,10 @@ #include "LLGI.CommandListDX12.h" #include "LLGI.PipelineStateDX12.h" #include "LLGI.PlatformDX12.h" +#include "LLGI.QueryDX12.h" #include "LLGI.ShaderDX12.h" #include "LLGI.SingleFrameMemoryPoolDX12.h" #include "LLGI.TextureDX12.h" -#include "LLGI.QueryDX12.h" namespace LLGI { @@ -310,6 +310,12 @@ std::vector GraphicsDX12::CaptureRenderTarget(Texture* renderTarget) auto dstFootprint = texture->GetFootprint().Footprint; + ID3D12CommandAllocator* commandAllocator = nullptr; + ID3D12GraphicsCommandList* commandList = nullptr; + D3D12_PLACED_SUBRESOURCE_FOOTPRINT footprint{}; + UINT64 totalSize{}; + D3D12_RESOURCE_DESC textureDesc{}; + BufferDX12 dstBuffer; if (!dstBuffer.Initialize(this, BufferUsageType::CopyDst | BufferUsageType::MapRead, dstFootprint.RowPitch * dstFootprint.Height)) { @@ -317,8 +323,6 @@ std::vector GraphicsDX12::CaptureRenderTarget(Texture* renderTarget) ::LLGI::Log(::LLGI::LogType::Error, msg.c_str()); goto FAILED_EXIT; } - ID3D12CommandAllocator* commandAllocator = nullptr; - ID3D12GraphicsCommandList* commandList = nullptr; auto hr = device->CreateCommandAllocator(D3D12_COMMAND_LIST_TYPE_DIRECT, IID_PPV_ARGS(&commandAllocator)); if (FAILED(hr)) @@ -339,9 +343,7 @@ std::vector GraphicsDX12::CaptureRenderTarget(Texture* renderTarget) goto FAILED_EXIT; } - D3D12_PLACED_SUBRESOURCE_FOOTPRINT footprint; - UINT64 totalSize; - auto textureDesc = texture->Get()->GetDesc(); + textureDesc = texture->Get()->GetDesc(); device->GetCopyableFootprints(&textureDesc, 0, 1, 0, &footprint, nullptr, nullptr, &totalSize); src.pResource = texture->Get(); @@ -374,11 +376,12 @@ std::vector GraphicsDX12::CaptureRenderTarget(Texture* renderTarget) if (GetTextureMemorySize(renderTarget->GetFormat(), rtSize3) != dstBuffer.GetSize()) { result.resize(GetTextureMemorySize(renderTarget->GetFormat(), rtSize3)); + const auto rowPitch = GetTextureRowPitch(renderTarget->GetFormat(), rtSize3); + const auto rowCount = GetTextureRowCount(renderTarget->GetFormat(), rtSize3); - for (int32_t y = 0; y < renderTarget->GetSizeAs2D().Y; y++) + for (int32_t y = 0; y < rowCount; y++) { - auto pitch = GetTextureMemorySize(renderTarget->GetFormat(), rtSize3) / renderTarget->GetSizeAs2D().Y; - memcpy(result.data() + pitch * y, raw + dstFootprint.RowPitch * y, pitch); + memcpy(result.data() + rowPitch * y, raw + dstFootprint.RowPitch * y, rowPitch); } } else @@ -408,9 +411,6 @@ Query* GraphicsDX12::CreateQuery(QueryType queryType, int32_t queryCount) return obj; } -uint64_t GraphicsDX12::TimestampToMicroseconds(uint64_t timestamp) const -{ - return timestamp * 1000000 / timestampFrequency_; -} +uint64_t GraphicsDX12::TimestampToMicroseconds(uint64_t timestamp) const { return timestamp * 1000000 / timestampFrequency_; } } // namespace LLGI diff --git a/src/DX12/LLGI.TextureDX12.cpp b/src/DX12/LLGI.TextureDX12.cpp index 69e11886..7f210635 100644 --- a/src/DX12/LLGI.TextureDX12.cpp +++ b/src/DX12/LLGI.TextureDX12.cpp @@ -195,7 +195,7 @@ void TextureDX12::CreateUploadReadbackBuffer() 1); assert(buffer_for_readback_ != nullptr); - if (static_cast(footprint_.Footprint.RowPitch) != cpu_memory_size_ / (texture_size_.Y * texture_size_.Z)) + if (static_cast(footprint_.Footprint.RowPitch) != GetTextureRowPitch(format_, texture_size_)) { locked_buffer_.resize(cpu_memory_size_); } @@ -222,11 +222,11 @@ void TextureDX12::Unlock() uint8_t* ptr = nullptr; buffer_for_upload_->Map(0, nullptr, (void**)&ptr); - int32_t rowCount = texture_size_.Y * texture_size_.Z; + const int32_t rowCount = GetTextureRowCount(format_, texture_size_); + const int32_t rowPitch = GetTextureRowPitch(format_, texture_size_); for (int32_t i = 0; i < rowCount; i++) { auto p = ptr + i * footprint_.Footprint.RowPitch; - auto rowPitch = cpu_memory_size_ / rowCount; memcpy(p, locked_buffer_.data() + rowPitch * i, rowPitch); } diff --git a/src/LLGI.Base.h b/src/LLGI.Base.h index f5bfa7cf..654b8d73 100644 --- a/src/LLGI.Base.h +++ b/src/LLGI.Base.h @@ -28,6 +28,7 @@ enum class DeviceType DirectX12, Metal, Vulkan, + WebGPU, }; enum class ErrorCode @@ -259,11 +260,13 @@ enum class TextureFormatType BC1, BC2, BC3, + BC7, R8G8B8A8_UNORM_SRGB, B8G8R8A8_UNORM_SRGB, BC1_SRGB, BC2_SRGB, BC3_SRGB, + BC7_SRGB, D32, D24S8, D32S8, @@ -509,6 +512,8 @@ inline std::string to_string(TextureFormatType format) return "BC2"; case TextureFormatType::BC3: return "BC3"; + case TextureFormatType::BC7: + return "BC7"; case TextureFormatType::R8G8B8A8_UNORM_SRGB: return "R8G8B8A8_UNORM_SRGB"; case TextureFormatType::B8G8R8A8_UNORM_SRGB: @@ -519,6 +524,8 @@ inline std::string to_string(TextureFormatType format) return "BC2_SRGB"; case TextureFormatType::BC3_SRGB: return "BC3_SRGB"; + case TextureFormatType::BC7_SRGB: + return "BC7_SRGB"; case TextureFormatType::D32: return "D32"; case TextureFormatType::D32S8: @@ -530,39 +537,109 @@ inline std::string to_string(TextureFormatType format) } } -inline int32_t GetTextureMemorySize(TextureFormatType format, Vec3I size) +inline bool IsBlockCompressedFormat(TextureFormatType format) { + switch (format) + { + case TextureFormatType::BC1: + return true; + case TextureFormatType::BC2: + return true; + case TextureFormatType::BC3: + return true; + case TextureFormatType::BC7: + return true; + case TextureFormatType::BC1_SRGB: + return true; + case TextureFormatType::BC2_SRGB: + return true; + case TextureFormatType::BC3_SRGB: + return true; + case TextureFormatType::BC7_SRGB: + return true; + default: + return false; + } +} + +inline int32_t GetTextureRowPitch(TextureFormatType format, Vec3I size) +{ + if (size.X <= 0 || size.Y <= 0 || size.Z <= 0) + return 0; + + const auto blockCountX = (size.X + 3) / 4; + switch (format) { case TextureFormatType::R8G8B8A8_UNORM: - return size.X * size.Y * size.Z * 4; + return size.X * 4; case TextureFormatType::B8G8R8A8_UNORM: - return size.X * size.Y * size.Z * 4; - case TextureFormatType::R8_UNORM: - return size.X * size.Y * size.Z * 1; + return size.X * 4; case TextureFormatType::R16G16_FLOAT: - return size.X * size.Y * size.Z * 4; - case TextureFormatType::R16G16B16A16_FLOAT: - return size.X * size.Y * size.Z * 8; - case TextureFormatType::R32G32B32A32_FLOAT: - return size.X * size.Y * size.Z * 16; + return size.X * 4; case TextureFormatType::R8G8B8A8_UNORM_SRGB: - return size.X * size.Y * size.Z * 4; + return size.X * 4; case TextureFormatType::B8G8R8A8_UNORM_SRGB: - return size.X * size.Y * size.Z * 4; + return size.X * 4; case TextureFormatType::D32: - return size.X * size.Y * size.Z * 4; + return size.X * 4; case TextureFormatType::D24S8: - return size.X * size.Y * size.Z * 4; + return size.X * 4; + case TextureFormatType::R8_UNORM: + return size.X; + case TextureFormatType::R16_FLOAT: + return size.X * 2; + case TextureFormatType::R32_FLOAT: + return size.X * 4; + case TextureFormatType::R32G32_FLOAT: + return size.X * 8; + case TextureFormatType::R16G16B16A16_FLOAT: + return size.X * 8; + case TextureFormatType::R32G32B32A32_FLOAT: + return size.X * 16; case TextureFormatType::D32S8: - return size.X * size.Y * size.Z * 5; + return size.X * 5; + case TextureFormatType::BC1: + return blockCountX * 8; + case TextureFormatType::BC1_SRGB: + return blockCountX * 8; + case TextureFormatType::BC2: + return blockCountX * 16; + case TextureFormatType::BC3: + return blockCountX * 16; + case TextureFormatType::BC7: + return blockCountX * 16; + case TextureFormatType::BC2_SRGB: + return blockCountX * 16; + case TextureFormatType::BC3_SRGB: + return blockCountX * 16; + case TextureFormatType::BC7_SRGB: + return blockCountX * 16; default: auto str = to_string(format); - Log(LogType::Error, str + " : GetTextureMemorySize is not supported"); + Log(LogType::Error, str + " : GetTextureRowPitch is not supported"); return 0; } } +inline int32_t GetTextureRowCount(TextureFormatType format, Vec3I size) +{ + if (size.Y <= 0 || size.Z <= 0) + return 0; + + if (IsBlockCompressedFormat(format)) + { + return ((size.Y + 3) / 4) * size.Z; + } + + return size.Y * size.Z; +} + +inline int32_t GetTextureMemorySize(TextureFormatType format, Vec3I size) +{ + return GetTextureRowPitch(format, size) * GetTextureRowCount(format, size); +} + inline uint32_t GetMaximumMipLevels(const Vec2I& size) { // (std::max) HACK for MSVC diff --git a/src/Mac/LLGI.WindowMac.mm b/src/Mac/LLGI.WindowMac.mm index 510e80b9..264024fb 100644 --- a/src/Mac/LLGI.WindowMac.mm +++ b/src/Mac/LLGI.WindowMac.mm @@ -165,7 +165,15 @@ bool newFrame() bool WindowMac::OnNewFrame() { return DoEvent(); } -void* WindowMac::GetNativePtr(int32_t index) { return GetNSWindowAsVoidPtr(); } +void* WindowMac::GetNativePtr(int32_t index) +{ + if (index == 1) + { + return impl_->window_.contentView; + } + + return GetNSWindowAsVoidPtr(); +} Vec2I WindowMac::GetWindowSize() const { return windowSize_; } diff --git a/src/Metal/LLGI.BufferMetal.mm b/src/Metal/LLGI.BufferMetal.mm index 1e71194c..3273f4e7 100644 --- a/src/Metal/LLGI.BufferMetal.mm +++ b/src/Metal/LLGI.BufferMetal.mm @@ -15,14 +15,15 @@ BufferMetal::BufferMetal() { - + } BufferMetal::~BufferMetal() { if (isExternalResource_) + { return; - + } if (buffer_ != nullptr) { [buffer_ release]; @@ -38,7 +39,7 @@ } auto g = static_cast(graphics); - + if(BitwiseContains(usage, BufferUsageType::MapWrite) || BitwiseContains(usage, BufferUsageType::MapRead)) { buffer_ = [g->GetDevice() newBufferWithLength:size options:MTLResourceStorageModeShared]; @@ -47,9 +48,9 @@ { buffer_ = [g->GetDevice() newBufferWithLength:size options:MTLResourceStorageModePrivate]; } - + size_ = size; - + return true; } diff --git a/src/Metal/LLGI.Metal_Impl.mm b/src/Metal/LLGI.Metal_Impl.mm index fc4888dc..a2984025 100644 --- a/src/Metal/LLGI.Metal_Impl.mm +++ b/src/Metal/LLGI.Metal_Impl.mm @@ -79,6 +79,14 @@ MTLPixelFormat ConvertFormat(TextureFormatType format) { return MTLPixelFormatBC3_RGBA_sRGB; } + else if (format == TextureFormatType::BC7) + { + return MTLPixelFormatBC7_RGBAUnorm; + } + else if (format == TextureFormatType::BC7_SRGB) + { + return MTLPixelFormatBC7_RGBAUnorm_sRGB; + } else if (format == TextureFormatType::D24S8) { return MTLPixelFormatDepth24Unorm_Stencil8; @@ -155,6 +163,14 @@ TextureFormatType ConvertFormat(MTLPixelFormat format) { return TextureFormatType::BC3_SRGB; } + else if (format == MTLPixelFormatBC7_RGBAUnorm) + { + return TextureFormatType::BC7; + } + else if (format == MTLPixelFormatBC7_RGBAUnorm_sRGB) + { + return TextureFormatType::BC7_SRGB; + } else if (format == MTLPixelFormatDepth24Unorm_Stencil8) { return TextureFormatType::D24S8; @@ -176,4 +192,4 @@ TextureFormatType ConvertFormat(MTLPixelFormat format) return TextureFormatType::Unknown; } -} +} // namespace LLGI diff --git a/src/PC/LLGI.CreatePC.cpp b/src/PC/LLGI.CreatePC.cpp index e0a72253..57e57e17 100644 --- a/src/PC/LLGI.CreatePC.cpp +++ b/src/PC/LLGI.CreatePC.cpp @@ -6,10 +6,12 @@ #include "../Vulkan/LLGI.PlatformVulkan.h" #endif -#ifdef _WIN32 +#if defined(_WIN32) && !defined(ENABLE_WEBGPU) #include "../DX12/LLGI.CompilerDX12.h" #include "../DX12/LLGI.PlatformDX12.h" #include "../Win/LLGI.WindowWin.h" +#elif defined(_WIN32) +#include "../Win/LLGI.WindowWin.h" #endif #ifdef __APPLE__ @@ -22,6 +24,11 @@ #include "../Vulkan/LLGI.CompilerVulkan.h" #endif +#ifdef ENABLE_WEBGPU +#include "../WebGPU/LLGI.CompilerWebGPU.h" +#include "../WebGPU/LLGI.PlatformWebGPU.h" +#endif + #ifdef __linux__ #include "../Linux/LLGI.WindowLinux.h" #endif @@ -33,6 +40,29 @@ namespace LLGI { +#ifdef __EMSCRIPTEN__ +namespace +{ +class WindowEmscripten : public Window +{ + Vec2I windowSize_; + +public: + explicit WindowEmscripten(Vec2I windowSize) : windowSize_(windowSize) {} + + bool OnNewFrame() override { return true; } + + void* GetNativePtr(int32_t index) override + { + (void)index; + return nullptr; + } + + Vec2I GetWindowSize() const override { return windowSize_; } +}; +} // namespace +#endif + Window* CreateWindow(const char* title, Vec2I windowSize) { #ifdef _WIN32 @@ -47,6 +77,9 @@ Window* CreateWindow(const char* title, Vec2I windowSize) { return window; } +#elif defined(__EMSCRIPTEN__) + (void)title; + return new WindowEmscripten(windowSize); #elif __linux__ auto window = new WindowLinux(); if (window->Initialize(title, windowSize)) @@ -65,6 +98,19 @@ Platform* CreatePlatform(const PlatformParameter& parameter, Window* window) windowSize.X = 1280; windowSize.Y = 720; +#ifdef ENABLE_WEBGPU + if (parameter.Device == DeviceType::WebGPU) + { + auto platform = new PlatformWebGPU(); + if (!platform->Initialize(window, parameter.WaitVSync)) + { + SafeRelease(platform); + return nullptr; + } + return platform; + } +#endif + #ifdef ENABLE_VULKAN #if defined(__linux__) if (parameter.Device == DeviceType::Vulkan || parameter.Device == DeviceType::Default) @@ -82,7 +128,7 @@ Platform* CreatePlatform(const PlatformParameter& parameter, Window* window) } #endif -#ifdef _WIN32 +#if defined(_WIN32) && !defined(ENABLE_WEBGPU) if (parameter.Device == DeviceType::Default || parameter.Device == DeviceType::DirectX12) { @@ -109,7 +155,7 @@ Compiler* CreateCompiler(DeviceType device) { #ifdef ENABLE_CREATE_COMPILER -#ifdef _WIN32 +#if defined(_WIN32) && !defined(ENABLE_WEBGPU) if (device == DeviceType::Default || device == DeviceType::DirectX12) { auto obj = new CompilerDX12(); @@ -129,6 +175,14 @@ Compiler* CreateCompiler(DeviceType device) } #endif +#ifdef ENABLE_WEBGPU + if (device == DeviceType::WebGPU) + { + auto obj = new CompilerWebGPU(); + return obj; + } +#endif + #ifdef __APPLE__ auto obj = new CompilerMetal(); return obj; diff --git a/src/Utils/LLGI.CommandListPool.h b/src/Utils/LLGI.CommandListPool.h index 696755d3..286db6d8 100644 --- a/src/Utils/LLGI.CommandListPool.h +++ b/src/Utils/LLGI.CommandListPool.h @@ -39,11 +39,19 @@ class CommandListPool SafeRelease(graphics_); } + void WaitUntilCompleted() + { + for (auto commandList : commandLists_) + { + commandList->WaitUntilCompleted(); + } + } + CommandList* Get(bool addRef = false) { CommandList* commandList = nullptr; - commandLists_[current_]->WaitUntilCompleted(); + WaitUntilCompleted(); commandList = commandLists_[current_]; diff --git a/src/Vulkan/LLGI.BaseVulkan.cpp b/src/Vulkan/LLGI.BaseVulkan.cpp index f89d4506..8e888592 100644 --- a/src/Vulkan/LLGI.BaseVulkan.cpp +++ b/src/Vulkan/LLGI.BaseVulkan.cpp @@ -1,5 +1,6 @@ #include "LLGI.BaseVulkan.h" #include "LLGI.GraphicsVulkan.h" +#include namespace LLGI { @@ -70,11 +71,13 @@ static FormatConversionItem s_formatConversionTable[] = { {TextureFormatType::BC1, VK_FORMAT_BC1_RGBA_UNORM_BLOCK}, {TextureFormatType::BC2, VK_FORMAT_BC2_UNORM_BLOCK}, {TextureFormatType::BC3, VK_FORMAT_BC3_UNORM_BLOCK}, + {TextureFormatType::BC7, VK_FORMAT_BC7_UNORM_BLOCK}, {TextureFormatType::R8G8B8A8_UNORM_SRGB, VK_FORMAT_R8G8B8A8_SRGB}, {TextureFormatType::B8G8R8A8_UNORM_SRGB, VK_FORMAT_B8G8R8A8_SRGB}, {TextureFormatType::BC1_SRGB, VK_FORMAT_BC1_RGBA_SRGB_BLOCK}, {TextureFormatType::BC2_SRGB, VK_FORMAT_BC2_SRGB_BLOCK}, {TextureFormatType::BC3_SRGB, VK_FORMAT_BC3_SRGB_BLOCK}, + {TextureFormatType::BC7_SRGB, VK_FORMAT_BC7_SRGB_BLOCK}, {TextureFormatType::D32, VK_FORMAT_D32_SFLOAT}, {TextureFormatType::D24S8, VK_FORMAT_D24_UNORM_S8_UINT}, {TextureFormatType::D32S8, VK_FORMAT_D32_SFLOAT_S8_UINT}, @@ -83,7 +86,7 @@ static FormatConversionItem s_formatConversionTable[] = { VkFormat VulkanHelper::TextureFormatToVkFormat(TextureFormatType format) { - for (size_t i = 0; i < sizeof(s_formatConversionTable); i++) + for (size_t i = 0; i < std::size(s_formatConversionTable); i++) { if (s_formatConversionTable[i].format == format) return s_formatConversionTable[i].vulkanFormat; @@ -97,7 +100,7 @@ VkFormat VulkanHelper::TextureFormatToVkFormat(TextureFormatType format) TextureFormatType VulkanHelper::VkFormatToTextureFormat(VkFormat format) { - for (size_t i = 0; i < sizeof(s_formatConversionTable); i++) + for (size_t i = 0; i < std::size(s_formatConversionTable); i++) { if (s_formatConversionTable[i].vulkanFormat == format) return s_formatConversionTable[i].format; diff --git a/src/Vulkan/LLGI.BaseVulkan.h b/src/Vulkan/LLGI.BaseVulkan.h index 22906875..84f75b5a 100644 --- a/src/Vulkan/LLGI.BaseVulkan.h +++ b/src/Vulkan/LLGI.BaseVulkan.h @@ -8,6 +8,10 @@ #ifdef _WIN32 #define VK_PROTOTYPES #define VK_USE_PLATFORM_WIN32_KHR +#elif defined(__APPLE__) +#define VK_PROTOTYPES +#define VK_ENABLE_BETA_EXTENSIONS +#define VK_USE_PLATFORM_MACOS_MVK #else #define VK_PROTOTYPES #define VK_USE_PLATFORM_XCB_KHR diff --git a/src/Vulkan/LLGI.GraphicsVulkan.cpp b/src/Vulkan/LLGI.GraphicsVulkan.cpp index 68215b5b..e205ac60 100644 --- a/src/Vulkan/LLGI.GraphicsVulkan.cpp +++ b/src/Vulkan/LLGI.GraphicsVulkan.cpp @@ -7,6 +7,7 @@ #include "LLGI.SingleFrameMemoryPoolVulkan.h" #include "LLGI.TextureVulkan.h" #include "LLGI.QueryVulkan.h" +#include "../LLGI.Platform.h" namespace LLGI { @@ -47,7 +48,16 @@ GraphicsVulkan::~GraphicsVulkan() SafeRelease(owner_); } -void GraphicsVulkan::SetWindowSize(const Vec2I& windowSize) { throw "Not inplemented"; } +void GraphicsVulkan::SetWindowSize(const Vec2I& windowSize) +{ + if (auto platform = dynamic_cast(owner_)) + { + platform->SetWindowSize(windowSize); + return; + } + + Graphics::SetWindowSize(windowSize); +} void GraphicsVulkan::Execute(CommandList* commandList) { diff --git a/src/Vulkan/LLGI.PipelineStateVulkan.cpp b/src/Vulkan/LLGI.PipelineStateVulkan.cpp index 296838fa..7634a274 100644 --- a/src/Vulkan/LLGI.PipelineStateVulkan.cpp +++ b/src/Vulkan/LLGI.PipelineStateVulkan.cpp @@ -400,7 +400,7 @@ bool PipelineStateVulkan::CreateGraphicsPipeline() blendFuncs[static_cast(BlendFuncType::DstColor)] = vk::BlendFactor::eDstColor; blendFuncs[static_cast(BlendFuncType::OneMinusDstColor)] = vk::BlendFactor::eOneMinusDstColor; blendFuncs[static_cast(BlendFuncType::DstAlpha)] = vk::BlendFactor::eDstAlpha; - blendFuncs[static_cast(BlendFuncType::OneMinusDstAlpha)] = vk::BlendFactor::eDstAlpha; + blendFuncs[static_cast(BlendFuncType::OneMinusDstAlpha)] = vk::BlendFactor::eOneMinusDstAlpha; blendInfo.srcColorBlendFactor = blendFuncs[static_cast(BlendSrcFunc)]; blendInfo.dstColorBlendFactor = blendFuncs[static_cast(BlendDstFunc)]; diff --git a/src/Vulkan/LLGI.PlatformVulkan.cpp b/src/Vulkan/LLGI.PlatformVulkan.cpp index 870bf154..5a2a7594 100644 --- a/src/Vulkan/LLGI.PlatformVulkan.cpp +++ b/src/Vulkan/LLGI.PlatformVulkan.cpp @@ -45,7 +45,8 @@ bool PlatformVulkan::CreateSwapChain(Vec2I windowSize, bool waitVSync) { auto oldSwapChain = swapchain_; - const auto disposeOldSwapchain = [&]() { + const auto disposeOldSwapchain = [&]() + { if (oldSwapChain) { for (uint32_t i = 0; i < swapBuffers.size(); i++) @@ -76,10 +77,13 @@ bool PlatformVulkan::CreateSwapChain(Vec2I windowSize, bool waitVSync) { disposeOldSwapchain(); swapchain_ = nullptr; + swapchainSize_ = {0, 0}; frameIndex = 0; } else { + swapchainSize_ = {static_cast(swapchainExtent.width), static_cast(swapchainExtent.height)}; + // select sync or vsync vk::PresentModeKHR swapchainPresentMode = vk::PresentModeKHR::eFifo; if (!waitVSync) @@ -158,7 +162,7 @@ bool PlatformVulkan::CreateSwapChain(Vec2I windowSize, bool waitVSync) swapBuffers[i].fence = vk::Fence(); swapBuffers[i].texture = new TextureVulkan(); - if (!swapBuffers[i].texture->InitializeAsScreen(swapBuffers[i].image, swapBuffers[i].view, surfaceFormat, windowSize)) + if (!swapBuffers[i].texture->InitializeAsScreen(swapBuffers[i].image, swapBuffers[i].view, surfaceFormat, swapchainSize_)) { Log(LogType::Error, "failed to create a texture while creating swap buffers."); throw "failed to create a texture while creating swap buffers."; @@ -196,6 +200,30 @@ bool PlatformVulkan::CreateDepthBuffer(Vec2I windowSize) return false; } +bool PlatformVulkan::RecreateSwapchain(const Vec2I& windowSize) +{ + if (!CreateSwapChain(windowSize, waitVSync_)) + { + return false; + } + + renderPasses_.clear(); + + if (!IsSwapchainValid()) + { + SafeRelease(depthStencilTexture_); + return true; + } + + if (!CreateDepthBuffer(swapchainSize_)) + { + return false; + } + + CreateRenderPass(); + return true; +} + void PlatformVulkan::CreateRenderPass() { renderPasses_.clear(); @@ -212,13 +240,16 @@ void PlatformVulkan::CreateRenderPass() } } -uint32_t PlatformVulkan::AcquireNextImage(vk::Semaphore& semaphore) +vk::Result PlatformVulkan::AcquireNextImage(vk::Semaphore& semaphore) { auto resultValue = vkDevice_.acquireNextImageKHR(swapchain_, UINT64_MAX, semaphore, vk::Fence()); - assert(resultValue.result == vk::Result::eSuccess); - frameIndex = resultValue.value; - return frameIndex; + if (resultValue.result == vk::Result::eSuccess || resultValue.result == vk::Result::eSuboptimalKHR) + { + frameIndex = resultValue.value; + } + + return resultValue.result; } vk::Fence PlatformVulkan::GetSubmitFence(bool destroy) @@ -459,6 +490,11 @@ bool PlatformVulkan::Initialize(Window* window, bool waitVSync) VK_KHR_SURFACE_EXTENSION_NAME, #ifdef _WIN32 VK_KHR_WIN32_SURFACE_EXTENSION_NAME, +#elif defined(__APPLE__) + VK_MVK_MACOS_SURFACE_EXTENSION_NAME, +#if defined(VK_KHR_PORTABILITY_ENUMERATION_EXTENSION_NAME) + VK_KHR_PORTABILITY_ENUMERATION_EXTENSION_NAME, +#endif #else VK_KHR_XCB_SURFACE_EXTENSION_NAME, #endif @@ -468,7 +504,8 @@ bool PlatformVulkan::Initialize(Window* window, bool waitVSync) #endif }; - auto exitWithError = [this]() -> void { + auto exitWithError = [this]() -> void + { Reset(); SafeRelease(depthStencilTexture_); @@ -493,6 +530,9 @@ bool PlatformVulkan::Initialize(Window* window, bool waitVSync) instanceCreateInfo.pApplicationInfo = &appInfo; instanceCreateInfo.enabledExtensionCount = static_cast(extensions.size()); instanceCreateInfo.ppEnabledExtensionNames = extensions.data(); +#if defined(__APPLE__) && defined(VK_INSTANCE_CREATE_ENUMERATE_PORTABILITY_BIT_KHR) + instanceCreateInfo.flags |= VK_INSTANCE_CREATE_ENUMERATE_PORTABILITY_BIT_KHR; +#endif #if !defined(NDEBUG) uint32_t layerCount = 0; @@ -519,6 +559,12 @@ bool PlatformVulkan::Initialize(Window* window, bool waitVSync) // get physics device auto physicalDevices = vkInstance_.enumeratePhysicalDevices(); + if (physicalDevices.empty()) + { + Log(LogType::Error, "No Vulkan physical devices found."); + exitWithError(); + return false; + } vkPhysicalDevice = physicalDevices[0]; struct Version @@ -539,6 +585,10 @@ bool PlatformVulkan::Initialize(Window* window, bool waitVSync) surfaceCreateInfo.hinstance = (HINSTANCE)window->GetNativePtr(1); surfaceCreateInfo.hwnd = (HWND)window->GetNativePtr(0); surface_ = vkInstance_.createWin32SurfaceKHR(surfaceCreateInfo); +#elif defined(__APPLE__) + vk::MacOSSurfaceCreateInfoMVK surfaceCreateInfo; + surfaceCreateInfo.pView = window->GetNativePtr(1); + surface_ = vkInstance_.createMacOSSurfaceMVK(surfaceCreateInfo); #else vk::XcbSurfaceCreateInfoKHR surfaceCreateInfo; surfaceCreateInfo.connection = XGetXCBConnection((Display*)window->GetNativePtr(0)); @@ -671,7 +721,7 @@ bool PlatformVulkan::Initialize(Window* window, bool waitVSync) vkCmdBuffers = vkDevice_.allocateCommandBuffers(allocInfo); // create depth buffer - if (!CreateDepthBuffer(window->GetWindowSize())) + if (IsSwapchainValid() && !CreateDepthBuffer(swapchainSize_)) { exitWithError(); return false; @@ -707,7 +757,28 @@ bool PlatformVulkan::NewFrame() if (IsSwapchainValid()) { - AcquireNextImage(vkPresentComplete_); + auto acquireResult = AcquireNextImage(vkPresentComplete_); + if (acquireResult == vk::Result::eErrorOutOfDateKHR) + { + vkDevice_.waitIdle(); + if (!RecreateSwapchain(windowSize_)) + { + return false; + } + + if (IsSwapchainValid()) + { + acquireResult = AcquireNextImage(vkPresentComplete_); + } + } + + if (IsSwapchainValid() && acquireResult != vk::Result::eSuccess && acquireResult != vk::Result::eSuboptimalKHR) + { + std::stringstream ss; + ss << "Failed to acquire next image : " << VulkanHelper::getResultName(static_cast(acquireResult)); + Log(LogType::Error, ss.str()); + return false; + } } executedCommandCount = 0; return true; @@ -764,12 +835,13 @@ void PlatformVulkan::Present() const auto result = Present(vkRenderComplete_); // TODO optimize it - if (result == vk::Result::eErrorOutOfDateKHR) + if (result == vk::Result::eErrorOutOfDateKHR || result == vk::Result::eSuboptimalKHR) { vkDevice_.waitIdle(); - CreateSwapChain(windowSize_, waitVSync_); - CreateDepthBuffer(windowSize_); - CreateRenderPass(); + if (!RecreateSwapchain(windowSize_)) + { + Log(LogType::Error, "Failed to recreate swapchain after present."); + } } } @@ -781,17 +853,18 @@ void PlatformVulkan::SetWindowSize(const Vec2I& windowSize) } vkDevice_.waitIdle(); - CreateSwapChain(windowSize, waitVSync_); - - CreateDepthBuffer(windowSize); - - CreateRenderPass(); + if (!RecreateSwapchain(windowSize)) + { + Log(LogType::Error, "Failed to recreate swapchain after resize."); + return; + } windowSize_ = windowSize; } Graphics* PlatformVulkan::CreateGraphics() { - auto addCommand = [this](vk::CommandBuffer commandBuffer, vk::Fence fence) -> void { + auto addCommand = [this](vk::CommandBuffer commandBuffer, vk::Fence fence) -> void + { std::array copySubmitInfos; copySubmitInfos[0].commandBufferCount = 1; copySubmitInfos[0].pCommandBuffers = &commandBuffer; diff --git a/src/Vulkan/LLGI.PlatformVulkan.h b/src/Vulkan/LLGI.PlatformVulkan.h index ed9b9ada..abfbe2aa 100644 --- a/src/Vulkan/LLGI.PlatformVulkan.h +++ b/src/Vulkan/LLGI.PlatformVulkan.h @@ -7,6 +7,8 @@ #ifdef _WIN32 #include "../Win/LLGI.WindowWin.h" +#elif defined(__APPLE__) +#include "../Mac/LLGI.WindowMac.h" #else #include "../Linux/LLGI.WindowLinux.h" #endif @@ -56,6 +58,7 @@ class PlatformVulkan : public Platform int32_t queueFamilyIndex_ = 0; Vec2I windowSize_; + Vec2I swapchainSize_; //! to check to finish present vk::Semaphore vkPresentComplete_; @@ -97,13 +100,15 @@ class PlatformVulkan : public Platform bool CreateDepthBuffer(Vec2I windowSize); + bool RecreateSwapchain(const Vec2I& windowSize); + void CreateRenderPass(); /*! @brief get swap buffer index @param semaphore the signaling semaphore to be waited for other functions */ - uint32_t AcquireNextImage(vk::Semaphore& semaphore); + vk::Result AcquireNextImage(vk::Semaphore& semaphore); vk::Fence GetSubmitFence(bool destroy = false); diff --git a/src/WebGPU/LLGI.BaseWebGPU.cpp b/src/WebGPU/LLGI.BaseWebGPU.cpp new file mode 100644 index 00000000..2e9a3417 --- /dev/null +++ b/src/WebGPU/LLGI.BaseWebGPU.cpp @@ -0,0 +1,338 @@ +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ + +wgpu::BlendOperation Convert(BlendEquationType type) +{ + if (type == BlendEquationType::Add) + return wgpu::BlendOperation::Add; + if (type == BlendEquationType::Max) + return wgpu::BlendOperation::Max; + if (type == BlendEquationType::Min) + return wgpu::BlendOperation::Min; + if (type == BlendEquationType::ReverseSub) + return wgpu::BlendOperation::ReverseSubtract; + if (type == BlendEquationType::Sub) + return wgpu::BlendOperation::Subtract; + + throw "Not implemented"; +} + +wgpu::BlendFactor Convert(BlendFuncType type) +{ + if (type == BlendFuncType::Zero) + return wgpu::BlendFactor::Zero; + if (type == BlendFuncType::One) + return wgpu::BlendFactor::One; + if (type == BlendFuncType::SrcColor) + return wgpu::BlendFactor::Src; + if (type == BlendFuncType::OneMinusSrcColor) + return wgpu::BlendFactor::OneMinusSrc; + if (type == BlendFuncType::SrcAlpha) + return wgpu::BlendFactor::SrcAlpha; + if (type == BlendFuncType::OneMinusSrcAlpha) + return wgpu::BlendFactor::OneMinusSrcAlpha; + if (type == BlendFuncType::DstAlpha) + return wgpu::BlendFactor::DstAlpha; + if (type == BlendFuncType::OneMinusDstAlpha) + return wgpu::BlendFactor::OneMinusDstAlpha; + if (type == BlendFuncType::DstColor) + return wgpu::BlendFactor::Dst; + if (type == BlendFuncType::OneMinusDstColor) + return wgpu::BlendFactor::OneMinusDst; + + throw "Not implemented"; +} + +wgpu::PrimitiveTopology Convert(TopologyType type) +{ + if (type == TopologyType::Point) + return wgpu::PrimitiveTopology::PointList; + + if (type == TopologyType::Line) + return wgpu::PrimitiveTopology::LineList; + + if (type == TopologyType::Triangle) + return wgpu::PrimitiveTopology::TriangleList; + + throw "Not implemented"; +} + +wgpu::CompareFunction Convert(CompareFuncType type) +{ + if (type == CompareFuncType::Always) + return wgpu::CompareFunction::Always; + + if (type == CompareFuncType::Equal) + return wgpu::CompareFunction::Equal; + + if (type == CompareFuncType::Greater) + return wgpu::CompareFunction::Greater; + + if (type == CompareFuncType::GreaterEqual) + return wgpu::CompareFunction::GreaterEqual; + + if (type == CompareFuncType::Less) + return wgpu::CompareFunction::Less; + + if (type == CompareFuncType::LessEqual) + return wgpu::CompareFunction::LessEqual; + + if (type == CompareFuncType::Never) + return wgpu::CompareFunction::Never; + + if (type == CompareFuncType::NotEqual) + return wgpu::CompareFunction::NotEqual; + + throw "Not implemented"; +} + +wgpu::CompareFunction Convert(DepthFuncType type) +{ + if (type == DepthFuncType::Always) + return wgpu::CompareFunction::Always; + + if (type == DepthFuncType::Equal) + return wgpu::CompareFunction::Equal; + + if (type == DepthFuncType::Greater) + return wgpu::CompareFunction::Greater; + + if (type == DepthFuncType::GreaterEqual) + return wgpu::CompareFunction::GreaterEqual; + + if (type == DepthFuncType::Less) + return wgpu::CompareFunction::Less; + + if (type == DepthFuncType::LessEqual) + return wgpu::CompareFunction::LessEqual; + + if (type == DepthFuncType::Never) + return wgpu::CompareFunction::Never; + + if (type == DepthFuncType::NotEqual) + return wgpu::CompareFunction::NotEqual; + + throw "Not implemented"; +} + + +wgpu::CullMode Convert(CullingMode mode) +{ + if (mode == CullingMode::Clockwise) + return wgpu::CullMode::Back; + + if (mode == CullingMode::CounterClockwise) + return wgpu::CullMode::Front; + + if (mode == CullingMode::DoubleSide) + return wgpu::CullMode::None; + + throw "Not implemented"; +} + +wgpu::VertexFormat Convert(VertexLayoutFormat format) +{ + if (format == VertexLayoutFormat::R32_FLOAT) + return wgpu::VertexFormat::Float32; + + if (format == VertexLayoutFormat::R32G32_FLOAT) + return wgpu::VertexFormat::Float32x2; + + if (format == VertexLayoutFormat::R32G32B32_FLOAT) + return wgpu::VertexFormat::Float32x3; + + if (format == VertexLayoutFormat::R32G32B32_FLOAT) + return wgpu::VertexFormat::Float32x3; + + if (format == VertexLayoutFormat::R8G8B8A8_UNORM) + return wgpu::VertexFormat::Unorm8x4; + + if (format == VertexLayoutFormat::R8G8B8A8_UINT) + return wgpu::VertexFormat::Uint8x4; + + throw "Not implemented"; +} + +wgpu::StencilOperation Convert(StencilOperatorType type) +{ + if (type == StencilOperatorType::Keep) + return wgpu::StencilOperation::Keep; + + if (type == StencilOperatorType::Zero) + return wgpu::StencilOperation::Zero; + + if (type == StencilOperatorType::Replace) + return wgpu::StencilOperation::Replace; + + if (type == StencilOperatorType::Invert) + return wgpu::StencilOperation::Invert; + + if (type == StencilOperatorType::IncClamp) + return wgpu::StencilOperation::IncrementClamp; + + if (type == StencilOperatorType::DecClamp) + return wgpu::StencilOperation::DecrementClamp; + + if (type == StencilOperatorType::IncRepeat) + return wgpu::StencilOperation::IncrementWrap; + + if (type == StencilOperatorType::DecRepeat) + return wgpu::StencilOperation::DecrementWrap; + + throw "Not implemented"; +} + +wgpu::TextureFormat ConvertFormat(TextureFormatType format) +{ + if (format == TextureFormatType::R8G8B8A8_UNORM) + return wgpu::TextureFormat::RGBA8Unorm; + + if (format == TextureFormatType::B8G8R8A8_UNORM) + return wgpu::TextureFormat::BGRA8Unorm; + + if (format == TextureFormatType::R16G16B16A16_FLOAT) + return wgpu::TextureFormat::RGBA16Float; + + if (format == TextureFormatType::R32G32B32A32_FLOAT) + return wgpu::TextureFormat::RGBA32Float; + + if (format == TextureFormatType::R8G8B8A8_UNORM_SRGB) + return wgpu::TextureFormat::RGBA8UnormSrgb; + + if (format == TextureFormatType::B8G8R8A8_UNORM_SRGB) + return wgpu::TextureFormat::BGRA8UnormSrgb; + + if (format == TextureFormatType::R16G16_FLOAT) + return wgpu::TextureFormat::RG16Float; + + if (format == TextureFormatType::R8_UNORM) + return wgpu::TextureFormat::R8Unorm; + + if (format == TextureFormatType::BC1) + return wgpu::TextureFormat::BC1RGBAUnorm; + + if (format == TextureFormatType::BC2) + return wgpu::TextureFormat::BC2RGBAUnorm; + + if (format == TextureFormatType::BC3) + return wgpu::TextureFormat::BC3RGBAUnorm; + + if (format == TextureFormatType::BC1_SRGB) + return wgpu::TextureFormat::BC1RGBAUnormSrgb; + + if (format == TextureFormatType::BC2_SRGB) + return wgpu::TextureFormat::BC2RGBAUnormSrgb; + + if (format == TextureFormatType::BC3_SRGB) + return wgpu::TextureFormat::BC3RGBAUnormSrgb; + + if (format == TextureFormatType::D32) + return wgpu::TextureFormat::Depth32Float; + + if (format == TextureFormatType::D24S8) + return wgpu::TextureFormat::Depth24PlusStencil8; + + if (format == TextureFormatType::D32S8) + return wgpu::TextureFormat::Depth32FloatStencil8; + + if (format == TextureFormatType::Unknown) + return wgpu::TextureFormat::Undefined; + + throw "Not implemented"; +} + +TextureFormatType ConvertFormat(wgpu::TextureFormat format) +{ + if (format == wgpu::TextureFormat::RGBA8Unorm) + return TextureFormatType::R8G8B8A8_UNORM; + + if (format == wgpu::TextureFormat::BGRA8Unorm) + return TextureFormatType::B8G8R8A8_UNORM; + + if (format == wgpu::TextureFormat::RGBA16Float) + return TextureFormatType::R16G16B16A16_FLOAT; + + if (format == wgpu::TextureFormat::RGBA32Float) + return TextureFormatType::R32G32B32A32_FLOAT; + + if (format == wgpu::TextureFormat::RGBA8UnormSrgb) + return TextureFormatType::R8G8B8A8_UNORM_SRGB; + + if (format == wgpu::TextureFormat::BGRA8UnormSrgb) + return TextureFormatType::B8G8R8A8_UNORM_SRGB; + + if (format == wgpu::TextureFormat::RG16Float) + return TextureFormatType::R16G16_FLOAT; + + if (format == wgpu::TextureFormat::R8Unorm) + return TextureFormatType::R8_UNORM; + + if (format == wgpu::TextureFormat::BC1RGBAUnorm) + return TextureFormatType::BC1; + + if (format == wgpu::TextureFormat::BC2RGBAUnorm) + return TextureFormatType::BC2; + + if (format == wgpu::TextureFormat::BC3RGBAUnorm) + return TextureFormatType::BC3; + + if (format == wgpu::TextureFormat::BC1RGBAUnormSrgb) + return TextureFormatType::BC1_SRGB; + + if (format == wgpu::TextureFormat::BC2RGBAUnormSrgb) + return TextureFormatType::BC2_SRGB; + + if (format == wgpu::TextureFormat::BC3RGBAUnormSrgb) + return TextureFormatType::BC3_SRGB; + + if (format == wgpu::TextureFormat::Depth32Float) + return TextureFormatType::D32; + + if (format == wgpu::TextureFormat::Depth24PlusStencil8) + return TextureFormatType::D24S8; + + if (format == wgpu::TextureFormat::Depth32FloatStencil8) + return TextureFormatType::D32S8; + + if (format == wgpu::TextureFormat::Undefined) + return TextureFormatType::Unknown; + + throw "Not implemented"; +} + +int32_t GetSize(VertexLayoutFormat format) +{ + if (format == VertexLayoutFormat::R32G32B32_FLOAT) + { + return sizeof(float) * 3; + } + else if (format == VertexLayoutFormat::R32G32B32A32_FLOAT) + { + return sizeof(float) * 4; + } + else if (format == VertexLayoutFormat::R32_FLOAT) + { + return sizeof(float) * 1; + } + else if (format == VertexLayoutFormat::R32G32_FLOAT) + { + return sizeof(float) * 2; + } + else if (format == VertexLayoutFormat::R8G8B8A8_UINT) + { + return sizeof(float); + } + else if (format == VertexLayoutFormat::R8G8B8A8_UNORM) + { + return sizeof(float); + } + else + { + Log(LogType::Error, "Unimplemented VertexLoayoutFormat"); + return 0; + } +} + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.BaseWebGPU.h b/src/WebGPU/LLGI.BaseWebGPU.h new file mode 100644 index 00000000..5b47d4f2 --- /dev/null +++ b/src/WebGPU/LLGI.BaseWebGPU.h @@ -0,0 +1,33 @@ + +#pragma once + +#include +#include +#include "../LLGI.Base.h" + +namespace LLGI +{ + +wgpu::BlendOperation Convert(BlendEquationType type); + +wgpu::BlendFactor Convert(BlendFuncType type); + +wgpu::PrimitiveTopology Convert(TopologyType type); + +wgpu::CompareFunction Convert(CompareFuncType type); + +wgpu::CompareFunction Convert(DepthFuncType type); + +wgpu::CullMode Convert(CullingMode mode); + +wgpu::VertexFormat Convert(VertexLayoutFormat format); + +wgpu::StencilOperation Convert(StencilOperatorType type); + +wgpu::TextureFormat ConvertFormat(TextureFormatType format); + +TextureFormatType ConvertFormat(wgpu::TextureFormat format); + +int32_t GetSize(VertexLayoutFormat format); + +} // namespace std diff --git a/src/WebGPU/LLGI.BufferWebGPU.cpp b/src/WebGPU/LLGI.BufferWebGPU.cpp new file mode 100644 index 00000000..64b60bd5 --- /dev/null +++ b/src/WebGPU/LLGI.BufferWebGPU.cpp @@ -0,0 +1,147 @@ +#include "LLGI.BufferWebGPU.h" + +#include +#include + +#if defined(__EMSCRIPTEN__) +#include +#endif + +namespace LLGI +{ + +namespace +{ +int32_t AlignTo(int32_t value, int32_t alignment) +{ + return (value + alignment - 1) / alignment * alignment; +} +} // namespace + +bool BufferWebGPU::Initialize(wgpu::Device& device, const BufferUsageType usage, const int32_t size, wgpu::Instance instance) +{ + device_ = device; + instance_ = instance; + + wgpu::BufferDescriptor desc{}; + allocatedSize_ = BitwiseContains(usage, BufferUsageType::Constant) ? AlignTo(size, 16) : size; + desc.size = allocatedSize_; + if (BitwiseContains(usage, BufferUsageType::MapRead)) + { + desc.usage = wgpu::BufferUsage::MapRead | wgpu::BufferUsage::CopyDst; + } + else + { + desc.usage = wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::CopySrc; + } + + if ((usage & BufferUsageType::Vertex) == BufferUsageType::Vertex) + { + desc.usage |= wgpu::BufferUsage::Vertex; + } + + if ((usage & BufferUsageType::Index) == BufferUsageType::Index) + { + desc.usage |= wgpu::BufferUsage::Index; + } + + if ((usage & BufferUsageType::Constant) == BufferUsageType::Constant) + { + desc.usage |= wgpu::BufferUsage::Uniform; + } + + if ((usage & BufferUsageType::ComputeRead) == BufferUsageType::ComputeRead || + (usage & BufferUsageType::ComputeWrite) == BufferUsageType::ComputeWrite) + { + desc.usage |= wgpu::BufferUsage::Storage; + } + + buffer_ = device.CreateBuffer(&desc); + size_ = size; + usage_ = usage; + return buffer_ != nullptr; +} + +void* BufferWebGPU::Lock() { return Lock(0, GetSize()); } + +void* BufferWebGPU::Lock(int32_t offset, int32_t size) +{ + lockedOffset_ = offset; + lockedSize_ = size; + + if (BitwiseContains(usage_, BufferUsageType::MapRead)) + { + bool completed = false; + bool succeeded = false; + auto future = buffer_.MapAsync(wgpu::MapMode::Read, + offset, + size, +#if defined(__EMSCRIPTEN__) + wgpu::CallbackMode::AllowSpontaneous, +#else + instance_ != nullptr ? wgpu::CallbackMode::WaitAnyOnly : wgpu::CallbackMode::AllowProcessEvents, +#endif + [&completed, &succeeded](wgpu::MapAsyncStatus status, wgpu::StringView) { + succeeded = status == wgpu::MapAsyncStatus::Success; + completed = true; + }); + + if (instance_ != nullptr) + { + instance_.WaitAny(future, 5ULL * 1000ULL * 1000ULL * 1000ULL); + } + else + { +#if defined(__EMSCRIPTEN__) + const double waitStart = emscripten_get_now(); + while (!completed) + { + emscripten_sleep(1); + if (emscripten_get_now() - waitStart > 5000.0) + { + break; + } + } +#else + const auto waitStart = std::chrono::steady_clock::now(); + while (!completed) + { + device_.Tick(); + if (std::chrono::steady_clock::now() - waitStart > std::chrono::seconds(5)) + { + break; + } + std::this_thread::sleep_for(std::chrono::milliseconds(1)); + } +#endif + } + + return succeeded ? const_cast(buffer_.GetConstMappedRange(offset, size)) : nullptr; + } + + lockedBuffer_.resize(size); + return lockedBuffer_.data(); +} + +void BufferWebGPU::Unlock() +{ + if (lockedBuffer_.empty()) + { + if (BitwiseContains(usage_, BufferUsageType::MapRead)) + { + buffer_.Unmap(); + } + return; + } + + device_.GetQueue().WriteBuffer(buffer_, lockedOffset_, lockedBuffer_.data(), lockedSize_); + lockedBuffer_.clear(); + lockedOffset_ = 0; + lockedSize_ = 0; +} + +int32_t BufferWebGPU::GetSize() { return size_; } + +wgpu::Buffer& BufferWebGPU::GetBuffer() { return buffer_; } + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.BufferWebGPU.h b/src/WebGPU/LLGI.BufferWebGPU.h new file mode 100644 index 00000000..7a4123ce --- /dev/null +++ b/src/WebGPU/LLGI.BufferWebGPU.h @@ -0,0 +1,37 @@ +#pragma once + +#include "../LLGI.Buffer.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ +/** + * TODO : Implement short time buffer +*/ +class BufferWebGPU : public Buffer +{ + wgpu::Buffer buffer_ = nullptr; + wgpu::Device device_ = nullptr; + wgpu::Instance instance_ = nullptr; + std::vector lockedBuffer_; + int32_t lockedOffset_ = 0; + int32_t lockedSize_ = 0; + int32_t size_ = 0; + int32_t allocatedSize_ = 0; + int32_t offset_ = 0; + +public: + bool Initialize(wgpu::Device& device, const BufferUsageType usage, const int32_t size, wgpu::Instance instance = nullptr); + void* Lock() override; + void* Lock(int32_t offset, int32_t size) override; + void Unlock() override; + + int32_t GetSize() override; + + int32_t GetOffset() const { return offset_; } + int32_t GetAllocatedSize() const { return allocatedSize_; } + + wgpu::Buffer& GetBuffer(); +}; + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.CommandListWebGPU.cpp b/src/WebGPU/LLGI.CommandListWebGPU.cpp new file mode 100644 index 00000000..76930975 --- /dev/null +++ b/src/WebGPU/LLGI.CommandListWebGPU.cpp @@ -0,0 +1,415 @@ +#include "LLGI.CommandListWebGPU.h" +#include "LLGI.BufferWebGPU.h" +#include "LLGI.PipelineStateWebGPU.h" +#include "LLGI.RenderPassWebGPU.h" +#include "LLGI.TextureWebGPU.h" + +#include + +namespace LLGI +{ + +CommandListWebGPU::CommandListWebGPU(wgpu::Device device) : device_(device) +{ + for (int w = 0; w < 2; w++) + { + for (int f = 0; f < 2; f++) + { + std::array filters; + filters[0] = wgpu::FilterMode::Nearest; + filters[1] = wgpu::FilterMode::Linear; + + std::array am; + am[0] = wgpu::AddressMode::ClampToEdge; + am[1] = wgpu::AddressMode::Repeat; + + wgpu::SamplerDescriptor samplerDesc; + + samplerDesc.magFilter = filters[f]; + samplerDesc.minFilter = filters[f]; + samplerDesc.maxAnisotropy = 1; + samplerDesc.addressModeU = am[w]; + samplerDesc.addressModeV = am[w]; + samplerDesc.addressModeW = am[w]; + samplers_[w][f] = device.CreateSampler(&samplerDesc); + } + } +} + +void CommandListWebGPU::Begin() +{ + wgpu::CommandEncoderDescriptor desc = {}; + commandEncorder_ = device_.CreateCommandEncoder(&desc); + + CommandList::Begin(); +} + +void CommandListWebGPU::End() +{ + commandBuffer_ = commandEncorder_.Finish(); + commandEncorder_ = nullptr; + + CommandList::End(); +} + +void CommandListWebGPU::BeginRenderPass(RenderPass* renderPass) +{ + auto rp = static_cast(renderPass); + rp->RefreshDescriptor(); + const auto& desc = rp->GetDescriptor(); + + renderPassEncorder_ = commandEncorder_.BeginRenderPass(&desc); + renderPassEncorder_.SetViewport(0.0f, 0.0f, static_cast(rp->GetScreenSize().X), static_cast(rp->GetScreenSize().Y), 0.0f, 1.0f); + + CommandList::BeginRenderPass(renderPass); +} + +void CommandListWebGPU::EndRenderPass() +{ + if (renderPassEncorder_ != nullptr) + { + renderPassEncorder_.End(); + renderPassEncorder_ = nullptr; + } + CommandList::EndRenderPass(); +} + +void CommandListWebGPU::BeginComputePass() +{ + wgpu::ComputePassDescriptor desc{}; + computePassEncorder_ = commandEncorder_.BeginComputePass(&desc); +} + +void CommandListWebGPU::EndComputePass() +{ + if (computePassEncorder_ != nullptr) + { + computePassEncorder_.End(); + computePassEncorder_ = nullptr; + } +} + +void CommandListWebGPU::Draw(int32_t primitiveCount, int32_t instanceCount) +{ + BindingVertexBuffer bvb; + BindingIndexBuffer bib; + PipelineState* bpip = nullptr; + + bool isVBDirtied = false; + bool isIBDirtied = false; + bool isPipDirtied = false; + + GetCurrentVertexBuffer(bvb, isVBDirtied); + GetCurrentIndexBuffer(bib, isIBDirtied); + GetCurrentPipelineState(bpip, isPipDirtied); + + assert(bvb.vertexBuffer != nullptr); + assert(bib.indexBuffer != nullptr); + assert(bpip != nullptr); + + auto vb = static_cast(bvb.vertexBuffer); + auto ib = static_cast(bib.indexBuffer); + auto pip = static_cast(bpip); + + if (vb != nullptr) + { + renderPassEncorder_.SetVertexBuffer(0, vb->GetBuffer(), bvb.offset, bvb.vertexBuffer->GetSize() - bvb.offset); + } + + if (ib != nullptr) + { + const auto format = bib.stride == 2 ? wgpu::IndexFormat::Uint16 : wgpu::IndexFormat::Uint32; + renderPassEncorder_.SetIndexBuffer(ib->GetBuffer(), format, bib.offset, ib->GetSize() - bib.offset); + } + + if (pip != nullptr) + { + renderPassEncorder_.SetPipeline(pip->GetRenderPipeline()); + renderPassEncorder_.SetStencilReference(pip->StencilRef); + } + + std::vector constantBindGroupEntries; + + for (size_t unit_ind = 0; unit_ind < constantBuffers_.size(); unit_ind++) + { + auto cb = static_cast(constantBuffers_[unit_ind]); + if (cb == nullptr) + { + continue; + } + + wgpu::BindGroupEntry entry = {}; + entry.binding = static_cast(unit_ind); + entry.buffer = cb->GetBuffer(); + entry.size = cb->GetAllocatedSize() - cb->GetOffset(); + entry.offset = cb->GetOffset(); + constantBindGroupEntries.push_back(entry); + } + + if (!constantBindGroupEntries.empty()) + { + wgpu::BindGroupDescriptor constantBindGroupDesc = {}; + constantBindGroupDesc.layout = pip->GetRenderPipeline().GetBindGroupLayout(0); + constantBindGroupDesc.entries = constantBindGroupEntries.data(); + constantBindGroupDesc.entryCount = constantBindGroupEntries.size(); + auto constantBindGroup = device_.CreateBindGroup(&constantBindGroupDesc); + renderPassEncorder_.SetBindGroup(0, constantBindGroup); + } + + std::vector textureGroupEntries; + std::vector samplerGroupEntries; + + for (int unit_ind = 0; unit_ind < static_cast(currentTextures_.size()); unit_ind++) + { + if (currentTextures_[unit_ind].texture == nullptr) + continue; + auto texture = static_cast(currentTextures_[unit_ind].texture); + auto wm = (int32_t)currentTextures_[unit_ind].wrapMode; + auto mm = (int32_t)currentTextures_[unit_ind].minMagFilter; + + wgpu::BindGroupEntry textureEntry = {}; + textureEntry.binding = unit_ind; + textureEntry.textureView = texture->GetTextureView(); + textureGroupEntries.push_back(textureEntry); + + wgpu::BindGroupEntry samplerEntry = {}; + if (!BitwiseContains(texture->GetParameter().Usage, TextureUsageType::Storage)) + { + samplerEntry.binding = unit_ind; + samplerEntry.sampler = samplers_[wm][mm]; + samplerGroupEntries.push_back(samplerEntry); + } + } + + for (int unit_ind = 0; unit_ind < static_cast(computeBuffers_.size()); unit_ind++) + { + if (computeBuffers_[unit_ind].computeBuffer == nullptr) + { + continue; + } + + auto buffer = static_cast(computeBuffers_[unit_ind].computeBuffer); + wgpu::BindGroupEntry bufferEntry = {}; + bufferEntry.binding = static_cast(unit_ind); + bufferEntry.buffer = buffer->GetBuffer(); + bufferEntry.offset = buffer->GetOffset(); + bufferEntry.size = buffer->GetSize(); + textureGroupEntries.push_back(bufferEntry); + } + + if (!textureGroupEntries.empty()) + { + wgpu::BindGroupDescriptor textureBindGroupDesc = {}; + textureBindGroupDesc.layout = pip->GetRenderPipeline().GetBindGroupLayout(1); + textureBindGroupDesc.entries = textureGroupEntries.data(); + textureBindGroupDesc.entryCount = textureGroupEntries.size(); + auto textureBindGroup = device_.CreateBindGroup(&textureBindGroupDesc); + renderPassEncorder_.SetBindGroup(1, textureBindGroup); + } + + if (!samplerGroupEntries.empty()) + { + wgpu::BindGroupDescriptor samplerBindGroupDesc = {}; + samplerBindGroupDesc.layout = pip->GetRenderPipeline().GetBindGroupLayout(2); + samplerBindGroupDesc.entries = samplerGroupEntries.data(); + samplerBindGroupDesc.entryCount = samplerGroupEntries.size(); + auto samplerBindGroup = device_.CreateBindGroup(&samplerBindGroupDesc); + renderPassEncorder_.SetBindGroup(2, samplerBindGroup); + } + + int indexPerPrim = 0; + + if (pip->Topology == TopologyType::Triangle) + { + indexPerPrim = 3; + } + else if (pip->Topology == TopologyType::Line) + { + indexPerPrim = 2; + } + else if (pip->Topology == TopologyType::Point) + { + indexPerPrim = 1; + } + else + { + assert(0); + } + + renderPassEncorder_.DrawIndexed(primitiveCount * indexPerPrim, instanceCount, 0, 0, 0); + CommandList::Draw(primitiveCount, instanceCount); +} + +void CommandListWebGPU::Dispatch(int32_t groupX, int32_t groupY, int32_t groupZ, int32_t threadX, int32_t threadY, int32_t threadZ) +{ + PipelineState* bpip = nullptr; + bool isPipDirtied = false; + GetCurrentPipelineState(bpip, isPipDirtied); + auto pip = static_cast(bpip); + if (pip == nullptr || computePassEncorder_ == nullptr) + { + return; + } + + computePassEncorder_.SetPipeline(pip->GetComputePipeline()); + + std::vector constantBindGroupEntries; + for (size_t unit_ind = 0; unit_ind < constantBuffers_.size(); unit_ind++) + { + auto cb = static_cast(constantBuffers_[unit_ind]); + if (cb == nullptr) + { + continue; + } + + wgpu::BindGroupEntry entry{}; + entry.binding = static_cast(unit_ind); + entry.buffer = cb->GetBuffer(); + entry.size = cb->GetAllocatedSize() - cb->GetOffset(); + entry.offset = cb->GetOffset(); + constantBindGroupEntries.push_back(entry); + } + + if (!constantBindGroupEntries.empty()) + { + wgpu::BindGroupDescriptor desc{}; + desc.layout = pip->GetComputePipeline().GetBindGroupLayout(0); + desc.entries = constantBindGroupEntries.data(); + desc.entryCount = constantBindGroupEntries.size(); + auto bindGroup = device_.CreateBindGroup(&desc); + computePassEncorder_.SetBindGroup(0, bindGroup); + } + + std::vector textureGroupEntries; + std::vector samplerAndBufferGroupEntries; + + for (int unit_ind = 0; unit_ind < static_cast(currentTextures_.size()); unit_ind++) + { + if (currentTextures_[unit_ind].texture == nullptr) + { + continue; + } + + auto texture = static_cast(currentTextures_[unit_ind].texture); + auto wm = (int32_t)currentTextures_[unit_ind].wrapMode; + auto mm = (int32_t)currentTextures_[unit_ind].minMagFilter; + + wgpu::BindGroupEntry textureEntry{}; + textureEntry.binding = unit_ind; + textureEntry.textureView = texture->GetTextureView(); + textureGroupEntries.push_back(textureEntry); + + if (!BitwiseContains(texture->GetParameter().Usage, TextureUsageType::Storage)) + { + wgpu::BindGroupEntry samplerEntry{}; + samplerEntry.binding = unit_ind; + samplerEntry.sampler = samplers_[wm][mm]; + samplerAndBufferGroupEntries.push_back(samplerEntry); + } + } + + for (int unit_ind = 0; unit_ind < static_cast(computeBuffers_.size()); unit_ind++) + { + if (computeBuffers_[unit_ind].computeBuffer == nullptr) + { + continue; + } + + auto buffer = static_cast(computeBuffers_[unit_ind].computeBuffer); + wgpu::BindGroupEntry entry{}; + entry.binding = static_cast(unit_ind); + entry.buffer = buffer->GetBuffer(); + entry.offset = buffer->GetOffset(); + entry.size = buffer->GetSize(); + samplerAndBufferGroupEntries.push_back(entry); + } + + if (!textureGroupEntries.empty()) + { + wgpu::BindGroupDescriptor desc{}; + desc.layout = pip->GetComputePipeline().GetBindGroupLayout(1); + desc.entries = textureGroupEntries.data(); + desc.entryCount = textureGroupEntries.size(); + auto bindGroup = device_.CreateBindGroup(&desc); + computePassEncorder_.SetBindGroup(1, bindGroup); + } + + if (!samplerAndBufferGroupEntries.empty()) + { + wgpu::BindGroupDescriptor desc{}; + desc.layout = pip->GetComputePipeline().GetBindGroupLayout(2); + desc.entries = samplerAndBufferGroupEntries.data(); + desc.entryCount = samplerAndBufferGroupEntries.size(); + auto bindGroup = device_.CreateBindGroup(&desc); + computePassEncorder_.SetBindGroup(2, bindGroup); + } + + computePassEncorder_.DispatchWorkgroups(groupX, groupY, groupZ); + CommandList::Dispatch(groupX, groupY, groupZ, threadX, threadY, threadZ); +} + +void CommandListWebGPU::SetScissor(int32_t x, int32_t y, int32_t width, int32_t height) +{ + renderPassEncorder_.SetScissorRect(x, y, width, height); +} + +void CommandListWebGPU::CopyTexture(Texture* src, Texture* dst) +{ + auto srcTex = static_cast(src); + CopyTexture(src, dst, {0, 0, 0}, {0, 0, 0}, srcTex->GetParameter().Size, 0, 0); +} + +void CommandListWebGPU::CopyTexture( + Texture* src, Texture* dst, const Vec3I& srcPos, const Vec3I& dstPos, const Vec3I& size, int srcLayer, int dstLayer) +{ + if (isInRenderPass_) + { + Log(LogType::Error, "Please call CopyTexture outside of RenderPass"); + return; + } + + auto srcTex = static_cast(src); + auto dstTex = static_cast(dst); + + wgpu::TexelCopyTextureInfo srcTexCopy; + wgpu::TexelCopyTextureInfo dstTexCopy; + wgpu::Extent3D extend3d; + + srcTexCopy.texture = srcTex->GetTexture(); + srcTexCopy.origin = {static_cast(srcPos.X), static_cast(srcPos.Y), static_cast(srcLayer + srcPos.Z)}; + srcTexCopy.aspect = wgpu::TextureAspect::All; + + dstTexCopy.texture = dstTex->GetTexture(); + dstTexCopy.origin = {static_cast(dstPos.X), static_cast(dstPos.Y), static_cast(dstLayer + dstPos.Z)}; + dstTexCopy.aspect = wgpu::TextureAspect::All; + + extend3d.width = size.X; + extend3d.height = size.Y; + extend3d.depthOrArrayLayers = size.Z; + + commandEncorder_.CopyTextureToTexture(&srcTexCopy, &dstTexCopy, &extend3d); +} + +void CommandListWebGPU::CopyBuffer(Buffer* src, Buffer* dst) +{ + auto srcBuffer = static_cast(src); + auto dstBuffer = static_cast(dst); + if (srcBuffer == nullptr || dstBuffer == nullptr) + { + return; + } + + commandEncorder_.CopyBufferToBuffer(srcBuffer->GetBuffer(), 0, dstBuffer->GetBuffer(), 0, std::min(srcBuffer->GetSize(), dstBuffer->GetSize())); +} + +void CommandListWebGPU::WaitUntilCompleted() +{ +#if !defined(__EMSCRIPTEN__) + if (device_ != nullptr) + { + device_.Tick(); + } +#endif +} + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.CommandListWebGPU.h b/src/WebGPU/LLGI.CommandListWebGPU.h new file mode 100644 index 00000000..64280e27 --- /dev/null +++ b/src/WebGPU/LLGI.CommandListWebGPU.h @@ -0,0 +1,51 @@ +#pragma once + +#include "../LLGI.CommandList.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ + +class CommandListWebGPU : public CommandList +{ + wgpu::Device device_; + wgpu::CommandBuffer commandBuffer_; + wgpu::CommandEncoder commandEncorder_; + wgpu::RenderPassEncoder renderPassEncorder_; + wgpu::ComputePassEncoder computePassEncorder_; + wgpu::Sampler samplers_[2][2]; + +public: + CommandListWebGPU(wgpu::Device device); + + void Begin() override; + + void End() override; + + void BeginRenderPass(RenderPass* renderPass) override; + + void EndRenderPass() override; + + void Draw(int32_t primitiveCount, int32_t instanceCount) override; + + void BeginComputePass() override; + + void EndComputePass() override; + + void Dispatch(int32_t groupX, int32_t groupY, int32_t groupZ, int32_t threadX, int32_t threadY, int32_t threadZ) override; + + void SetScissor(int32_t x, int32_t y, int32_t width, int32_t height) override; + + void CopyTexture(Texture* src, Texture* dst) override; + + void CopyTexture( + Texture* src, Texture* dst, const Vec3I& srcPos, const Vec3I& dstPos, const Vec3I& size, int srcLayer, int dstLayer) override; + + void CopyBuffer(Buffer* src, Buffer* dst) override; + + void WaitUntilCompleted() override; + + const wgpu::CommandBuffer& GetCommandBuffer() const { return commandBuffer_; } +}; + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.CompilerWebGPU.cpp b/src/WebGPU/LLGI.CompilerWebGPU.cpp new file mode 100644 index 00000000..6858f9ae --- /dev/null +++ b/src/WebGPU/LLGI.CompilerWebGPU.cpp @@ -0,0 +1,88 @@ +#include "LLGI.CompilerWebGPU.h" +#include + +namespace LLGI +{ +void CompilerWebGPU::Compile(CompilerResult& result, const char* code, ShaderStageType shaderStage) +{ + std::vector buffer; + + // header + buffer.push_back('w'); + buffer.push_back('g'); + buffer.push_back('s'); + buffer.push_back('l'); + buffer.push_back('c'); + buffer.push_back('o'); + buffer.push_back('d'); + buffer.push_back('e'); + + const auto len = strlen(code) + 1; + for (size_t i = 0; i < len; i++) + { + buffer.push_back(code[i]); + } + + result.Binary.resize(1); + result.Binary[0].resize(buffer.size()); + memcpy(result.Binary[0].data(), buffer.data(), buffer.size()); +} +} // namespace LLGI + +/* +#include "LLGI.CompilerMetal.h" + +#import + +namespace LLGI +{ + +void CompilerMetal::Initialize() {} + +void CompilerMetal::Compile(CompilerResult& result, const char* code, ShaderStageType shaderStage) +{ + @autoreleasepool + { + // Metal doesn't support to save a library as binary file (with external tool, it can) + NSString* codeStr = [[[NSString alloc] initWithUTF8String:code] autorelease]; + + id device = MTLCreateSystemDefaultDevice(); + + NSError* libraryError = nil; + id lib = [[device newLibraryWithSource:codeStr options:NULL error:&libraryError] autorelease]; + if (libraryError) + { + result.Message = libraryError.localizedDescription.UTF8String; + } + + if (lib == NULL) + { + return; + } + + std::vector buffer; + + // header + buffer.push_back('m'); + buffer.push_back('t'); + buffer.push_back('l'); + buffer.push_back('c'); + buffer.push_back('o'); + buffer.push_back('d'); + buffer.push_back('e'); + + auto len = strlen(code) + 1; + for (int i = 0; i < len; i++) + { + buffer.push_back(code[i]); + } + + result.Binary.resize(1); + result.Binary[0].resize(buffer.size()); + memcpy(result.Binary[0].data(), buffer.data(), buffer.size()); + } +} + +} + +*/ \ No newline at end of file diff --git a/src/WebGPU/LLGI.CompilerWebGPU.h b/src/WebGPU/LLGI.CompilerWebGPU.h new file mode 100644 index 00000000..fac97434 --- /dev/null +++ b/src/WebGPU/LLGI.CompilerWebGPU.h @@ -0,0 +1,17 @@ +#pragma once + +#include "LLGI.BaseWebGPU.h" +#include "../LLGI.Compiler.h" + +namespace LLGI +{ + +class CompilerWebGPU : public Compiler +{ +public: + void Compile(CompilerResult& result, const char* code, ShaderStageType shaderStage) override; + + DeviceType GetDeviceType() const override { return DeviceType::WebGPU; } +}; + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.GraphicsWebGPU.cpp b/src/WebGPU/LLGI.GraphicsWebGPU.cpp new file mode 100644 index 00000000..5efbf08d --- /dev/null +++ b/src/WebGPU/LLGI.GraphicsWebGPU.cpp @@ -0,0 +1,416 @@ +#include "LLGI.GraphicsWebGPU.h" + +#include "LLGI.BufferWebGPU.h" +#include "LLGI.CommandListWebGPU.h" +#include "LLGI.PipelineStateWebGPU.h" +#include "LLGI.RenderPassPipelineStateWebGPU.h" +#include "LLGI.RenderPassWebGPU.h" +#include "LLGI.ShaderWebGPU.h" +#include "LLGI.TextureWebGPU.h" + +#include +#include +#include +#include + +#if defined(__EMSCRIPTEN__) +#include +#endif + +namespace LLGI +{ + +namespace +{ +uint32_t AlignTo(uint32_t value, uint32_t alignment) +{ + return (value + alignment - 1) / alignment * alignment; +} + +uint32_t GetFormatBytesPerPixel(TextureFormatType format) +{ + switch (format) + { + case TextureFormatType::R8_UNORM: + return 1; + case TextureFormatType::R32G32B32A32_FLOAT: + return 16; + default: + return 4; + } +} + +#if defined(__EMSCRIPTEN__) +void WaitForQueue(wgpu::Queue& queue) +{ + bool completed = false; + bool succeeded = false; + queue.OnSubmittedWorkDone(wgpu::CallbackMode::AllowSpontaneous, + [&completed, &succeeded](wgpu::QueueWorkDoneStatus status, wgpu::StringView) { + succeeded = status == wgpu::QueueWorkDoneStatus::Success; + completed = true; + }); + + const double waitStart = emscripten_get_now(); + while (!completed) + { + emscripten_sleep(1); + if (emscripten_get_now() - waitStart > 5000.0) + { + break; + } + } + + if (!succeeded) + { + Log(LogType::Warning, "Timed out or failed while waiting for WebGPU queue completion."); + } +} +#endif +} // namespace + +class SingleFrameMemoryPoolWebGPU : public SingleFrameMemoryPool +{ + wgpu::Device device_; + + Buffer* CreateBufferInternal(int32_t size) override + { + auto obj = new BufferWebGPU(); + if (!obj->Initialize(device_, BufferUsageType::Constant, size)) + { + SafeRelease(obj); + return nullptr; + } + return obj; + } + + Buffer* ReinitializeBuffer(Buffer* cb, int32_t size) override + { + if (cb != nullptr && cb->GetSize() >= size) + { + return cb; + } + return CreateBufferInternal(size); + } + +public: + SingleFrameMemoryPoolWebGPU(wgpu::Device device, int32_t swapBufferCount) + : SingleFrameMemoryPool(swapBufferCount), device_(device) + { + } +}; + +GraphicsWebGPU::GraphicsWebGPU(wgpu::Device device) : device_(device) { queue_ = device.GetQueue(); } + +GraphicsWebGPU::GraphicsWebGPU(wgpu::Device device, wgpu::Instance instance) : device_(device), instance_(instance) +{ + queue_ = device.GetQueue(); +} + +void GraphicsWebGPU::SetWindowSize(const Vec2I& windowSize) {} + +void GraphicsWebGPU::Execute(CommandList* commandList) +{ + auto commandListWgpu = static_cast(commandList); + auto cb = commandListWgpu->GetCommandBuffer(); + queue_.Submit(1, &cb); +} + +void GraphicsWebGPU::WaitFinish() +{ +#if defined(__EMSCRIPTEN__) + if (queue_ != nullptr) + { + WaitForQueue(queue_); + } +#else + if (device_ != nullptr) + { + device_.Tick(); + } +#endif +} + +Buffer* GraphicsWebGPU::CreateBuffer(BufferUsageType usage, int32_t size) +{ + auto obj = new BufferWebGPU(); + if (!obj->Initialize(GetDevice(), usage, size, instance_)) + { + SafeRelease(obj); + return nullptr; + } + + return obj; +} + +Shader* GraphicsWebGPU::CreateShader(DataStructure* data, int32_t count) +{ + auto obj = new ShaderWebGPU(); + if (!obj->Initialize(GetDevice(), data, count)) + { + SafeRelease(obj); + return nullptr; + } + return obj; +} + +PipelineState* GraphicsWebGPU::CreatePiplineState() +{ + auto pipelineState = new PipelineStateWebGPU(GetDevice()); + + // TODO : error check + return pipelineState; +} + +SingleFrameMemoryPool* GraphicsWebGPU::CreateSingleFrameMemoryPool(int32_t constantBufferPoolSize, int32_t drawingCount) +{ + return new SingleFrameMemoryPoolWebGPU(GetDevice(), 1); +} + +CommandList* GraphicsWebGPU::CreateCommandList(SingleFrameMemoryPool* memoryPool) +{ + auto commandList = new CommandListWebGPU(GetDevice()); + + // TODO : error check + return commandList; +} + +RenderPass* GraphicsWebGPU::CreateRenderPass(Texture** textures, int32_t textureCount, Texture* depthTexture) +{ + assert(textures != nullptr); + if (textures == nullptr) + return nullptr; + + for (int32_t i = 0; i < textureCount; i++) + { + assert(textures[i] != nullptr); + if (textures[i] == nullptr) + return nullptr; + } + + auto dt = static_cast(depthTexture); + + auto renderPass = new RenderPassWebGPU(); + if (!renderPass->Initialize(textures, textureCount, dt, nullptr, nullptr)) + { + SafeRelease(renderPass); + } + + return renderPass; +} + +RenderPass* +GraphicsWebGPU::CreateRenderPass(Texture* texture, Texture* resolvedTexture, Texture* depthTexture, Texture* resolvedDepthTexture) +{ + if (texture == nullptr) + return nullptr; + + auto dt = static_cast(depthTexture); + auto rt = static_cast(resolvedTexture); + auto rdt = static_cast(resolvedDepthTexture); + + auto renderPass = new RenderPassWebGPU(); + if (!renderPass->Initialize((&texture), 1, (TextureWebGPU*)dt, (TextureWebGPU*)rt, (TextureWebGPU*)rdt)) + { + SafeRelease(renderPass); + } + + return renderPass; +} + +Texture* GraphicsWebGPU::CreateTexture(uint64_t id) { return nullptr; } + +Texture* GraphicsWebGPU::CreateTexture(const TextureParameter& parameter) +{ + auto obj = new TextureWebGPU(); + if (!obj->Initialize(GetDevice(), parameter, instance_)) + { + SafeRelease(obj); + return nullptr; + } + return obj; +} + +Texture* GraphicsWebGPU::CreateTexture(const TextureInitializationParameter& parameter) +{ + TextureParameter param; + param.Dimension = 2; + param.Format = parameter.Format; + param.MipLevelCount = parameter.MipMapCount; + param.SampleCount = 1; + param.Size = {parameter.Size.X, parameter.Size.Y, 1}; + return CreateTexture(param); +} + +Texture* GraphicsWebGPU::CreateRenderTexture(const RenderTextureInitializationParameter& parameter) +{ + TextureParameter param; + param.Dimension = 2; + param.Format = parameter.Format; + param.MipLevelCount = 1; + param.SampleCount = parameter.SamplingCount; + param.Size = {parameter.Size.X, parameter.Size.Y, 1}; + param.Usage = TextureUsageType::RenderTarget; + return CreateTexture(param); +} + +Texture* GraphicsWebGPU::CreateDepthTexture(const DepthTextureInitializationParameter& parameter) +{ + auto format = TextureFormatType::D32; + if (parameter.Mode == DepthTextureMode::DepthStencil) + { + format = TextureFormatType::D24S8; + } + + TextureParameter param; + param.Dimension = 2; + param.Format = format; + param.MipLevelCount = 1; + param.SampleCount = parameter.SamplingCount; + param.Size = {parameter.Size.X, parameter.Size.Y, 1}; + param.Usage = TextureUsageType::RenderTarget; + return CreateTexture(param); +} + +std::vector GraphicsWebGPU::CaptureRenderTarget(Texture* renderTarget) +{ + auto texture = static_cast(renderTarget); + if (texture == nullptr) + { + return std::vector(); + } + + const auto size = texture->GetSizeAs2D(); + const auto bytesPerPixel = GetFormatBytesPerPixel(texture->GetFormat()); + const auto unalignedBytesPerRow = static_cast(size.X) * bytesPerPixel; + const auto bytesPerRow = AlignTo(unalignedBytesPerRow, 256); + const auto bufferSize = static_cast(bytesPerRow) * static_cast(size.Y); + + wgpu::BufferDescriptor bufferDesc{}; + bufferDesc.size = bufferSize; + bufferDesc.usage = wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::MapRead; + auto readbackBuffer = device_.CreateBuffer(&bufferDesc); + + wgpu::CommandEncoderDescriptor encoderDesc{}; + auto encoder = device_.CreateCommandEncoder(&encoderDesc); + + wgpu::TexelCopyTextureInfo src{}; + src.texture = texture->GetTexture(); + src.aspect = wgpu::TextureAspect::All; + + wgpu::TexelCopyBufferInfo dst{}; + dst.buffer = readbackBuffer; + dst.layout.offset = 0; + dst.layout.bytesPerRow = bytesPerRow; + dst.layout.rowsPerImage = static_cast(size.Y); + + wgpu::Extent3D extent{}; + extent.width = static_cast(size.X); + extent.height = static_cast(size.Y); + extent.depthOrArrayLayers = 1; + encoder.CopyTextureToBuffer(&src, &dst, &extent); + + auto commandBuffer = encoder.Finish(); + queue_.Submit(1, &commandBuffer); + WaitFinish(); + + bool completed = false; + bool succeeded = false; + auto future = readbackBuffer.MapAsync(wgpu::MapMode::Read, + 0, + bufferSize, +#if defined(__EMSCRIPTEN__) + wgpu::CallbackMode::AllowSpontaneous, +#else + instance_ != nullptr ? wgpu::CallbackMode::WaitAnyOnly : wgpu::CallbackMode::AllowProcessEvents, +#endif + [&completed, &succeeded](wgpu::MapAsyncStatus status, wgpu::StringView) { + succeeded = status == wgpu::MapAsyncStatus::Success; + completed = true; + }); + + if (instance_ != nullptr) + { + instance_.WaitAny(future, 5ULL * 1000ULL * 1000ULL * 1000ULL); + } + else + { +#if defined(__EMSCRIPTEN__) + const double waitStart = emscripten_get_now(); + while (!completed) + { + emscripten_sleep(1); + if (emscripten_get_now() - waitStart > 5000.0) + { + break; + } + } +#else + const auto waitStart = std::chrono::steady_clock::now(); + while (!completed) + { + device_.Tick(); + if (std::chrono::steady_clock::now() - waitStart > std::chrono::seconds(5)) + { + break; + } + std::this_thread::sleep_for(std::chrono::milliseconds(1)); + } +#endif + } + + std::vector ret(static_cast(unalignedBytesPerRow) * static_cast(size.Y)); + if (!succeeded) + { + Log(LogType::Warning, "Timed out or failed while waiting for WebGPU readback."); + return ret; + } + + auto mapped = static_cast(readbackBuffer.GetConstMappedRange(0, bufferSize)); + for (int32_t y = 0; y < size.Y; y++) + { + memcpy(ret.data() + static_cast(y) * unalignedBytesPerRow, + mapped + static_cast(y) * bytesPerRow, + unalignedBytesPerRow); + } + readbackBuffer.Unmap(); + return ret; +} + +RenderPassPipelineState* GraphicsWebGPU::CreateRenderPassPipelineState(RenderPass* renderPass) +{ + return CreateRenderPassPipelineState(renderPass->GetKey()); +} + +RenderPassPipelineState* GraphicsWebGPU::CreateRenderPassPipelineState(const RenderPassPipelineStateKey& key) +{ + // already? + { + auto it = renderPassPipelineStates_.find(key); + + if (it != renderPassPipelineStates_.end()) + { + auto ret = it->second; + + if (ret != nullptr) + { + auto ptr = ret.get(); + SafeAddRef(ptr); + return ptr; + } + } + } + + std::shared_ptr ret = LLGI::CreateSharedPtr<>(new RenderPassPipelineStateWebGPU()); + ret->SetKey(key); + + renderPassPipelineStates_[key] = ret; + + { + auto ptr = ret.get(); + SafeAddRef(ptr); + return ptr; + } +} + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.GraphicsWebGPU.h b/src/WebGPU/LLGI.GraphicsWebGPU.h new file mode 100644 index 00000000..f5c638e8 --- /dev/null +++ b/src/WebGPU/LLGI.GraphicsWebGPU.h @@ -0,0 +1,67 @@ +#pragma once + +#include "../LLGI.Base.h" +#include "../LLGI.Graphics.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ +class RenderPassPipelineStateWebGPU; + +class GraphicsWebGPU : public Graphics +{ +private: + //! cached + std::unordered_map, RenderPassPipelineStateKey::Hash> + renderPassPipelineStates_; + + wgpu::Device device_; + wgpu::Instance instance_; + wgpu::Queue queue_; + +public: + + GraphicsWebGPU(wgpu::Device device); + GraphicsWebGPU(wgpu::Device device, wgpu::Instance instance); + + void SetWindowSize(const Vec2I& windowSize) override; + + void Execute(CommandList* commandList) override; + + void WaitFinish() override; + + Buffer* CreateBuffer(BufferUsageType usage, int32_t size) override; + + Shader* CreateShader(DataStructure* data, int32_t count) override; + + PipelineState* CreatePiplineState() override; + + SingleFrameMemoryPool* CreateSingleFrameMemoryPool(int32_t constantBufferPoolSize, int32_t drawingCount) override; + + CommandList* CreateCommandList(SingleFrameMemoryPool* memoryPool) override; + + RenderPass* CreateRenderPass(Texture** textures, int32_t textureCount, Texture* depthTexture) override; + + RenderPass* CreateRenderPass(Texture* texture, Texture* resolvedTexture, Texture* depthTexture, Texture* resolvedDepthTexture) override; + + Texture* CreateTexture(const TextureParameter& parameter) override; + + Texture* CreateTexture(const TextureInitializationParameter& parameter) override; + + Texture* CreateRenderTexture(const RenderTextureInitializationParameter& parameter) override; + + Texture* CreateDepthTexture(const DepthTextureInitializationParameter& parameter) override; + + Texture* CreateTexture(uint64_t id) override; + + std::vector CaptureRenderTarget(Texture* renderTarget) override; + + RenderPassPipelineState* CreateRenderPassPipelineState(RenderPass* renderPass) override; + + RenderPassPipelineState* CreateRenderPassPipelineState(const RenderPassPipelineStateKey& key) override; + + wgpu::Device& GetDevice() { return device_; } + wgpu::Queue& GetQueue() { return queue_; } +}; + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.PipelineStateWebGPU.cpp b/src/WebGPU/LLGI.PipelineStateWebGPU.cpp new file mode 100644 index 00000000..7483f684 --- /dev/null +++ b/src/WebGPU/LLGI.PipelineStateWebGPU.cpp @@ -0,0 +1,159 @@ +#include "LLGI.PipelineStateWebGPU.h" +#include "LLGI.RenderPassPipelineStateWebGPU.h" +#include "LLGI.ShaderWebGPU.h" +#include + +namespace LLGI +{ + +PipelineStateWebGPU::PipelineStateWebGPU(wgpu::Device device) : device_(device) { shaders_.fill(nullptr); } + +PipelineStateWebGPU::~PipelineStateWebGPU() +{ + for (auto& shader : shaders_) + { + SafeRelease(shader); + } +} + +void PipelineStateWebGPU::SetShader(ShaderStageType stage, Shader* shader) +{ + SafeAddRef(shader); + SafeRelease(shaders_[static_cast(stage)]); + shaders_[static_cast(stage)] = shader; +} + +bool PipelineStateWebGPU::Compile() +{ + const char* entryPointName = "main"; + auto computeShader = static_cast(shaders_[static_cast(ShaderStageType::Compute)]); + if (computeShader != nullptr) + { + wgpu::ComputePipelineDescriptor desc{}; + desc.layout = nullptr; + desc.compute.module = computeShader->GetShaderModule(); + desc.compute.entryPoint = entryPointName; + computePipeline_ = device_.CreateComputePipeline(&desc); + return computePipeline_ != nullptr; + } + + wgpu::RenderPipelineDescriptor desc{}; + + desc.primitive.topology = Convert(Topology); + desc.primitive.stripIndexFormat = wgpu::IndexFormat::Undefined; // is it correct? + desc.primitive.frontFace = wgpu::FrontFace::CW; + desc.primitive.cullMode = Convert(Culling); + desc.multisample.count = static_cast(renderPassPipelineState_->Key.SamplingCount); + desc.multisample.mask = std::numeric_limits::max(); + desc.multisample.alphaToCoverageEnabled = false; + desc.layout = nullptr; // is it correct? + + auto vertexShader = static_cast(shaders_[static_cast(ShaderStageType::Vertex)]); + + desc.vertex.module = vertexShader->GetShaderModule(); + desc.vertex.entryPoint = entryPointName; + + desc.vertex.bufferCount = 1; + std::array bufferLayouts; + desc.vertex.buffers = bufferLayouts.data(); + + bufferLayouts[0].attributeCount = VertexLayoutCount; + bufferLayouts[0].arrayStride = 0; + bufferLayouts[0].stepMode = wgpu::VertexStepMode::Vertex; + + std::array attributes; + bufferLayouts[0].attributes = attributes.data(); + + int offset = 0; + for (int i = 0; i < VertexLayoutCount; i++) + { + attributes[i].format = Convert(VertexLayouts[i]); + attributes[i].offset = offset; + attributes[i].shaderLocation = i; + offset += GetSize(VertexLayouts[i]); + } + bufferLayouts[0].arrayStride = offset; + + auto pixelShader = static_cast(shaders_[static_cast(ShaderStageType::Pixel)]); + + // TODO : support blend enabled + wgpu::BlendState blendState; + blendState.color.srcFactor = Convert(BlendSrcFunc); + blendState.color.dstFactor = Convert(BlendDstFunc); + blendState.color.operation = Convert(BlendEquationRGB); + blendState.alpha.srcFactor = Convert(BlendSrcFuncAlpha); + blendState.alpha.dstFactor = Convert(BlendDstFuncAlpha); + blendState.alpha.operation = Convert(BlendEquationAlpha); + + std::array colorTargetStates; + + for (size_t i = 0; i < renderPassPipelineState_->Key.RenderTargetFormats.size(); i++) + { + colorTargetStates[i].blend = IsBlendEnabled ? &blendState : nullptr; + colorTargetStates[i].format = ConvertFormat(renderPassPipelineState_->Key.RenderTargetFormats.at(i)); + colorTargetStates[i].writeMask = wgpu::ColorWriteMask::All; + } + + wgpu::FragmentState fragmentState = {}; + fragmentState.targetCount = static_cast(renderPassPipelineState_->Key.RenderTargetFormats.size()); + fragmentState.targets = colorTargetStates.data(); + fragmentState.entryPoint = entryPointName; + fragmentState.module = pixelShader->GetShaderModule(); + + desc.fragment = &fragmentState; + + wgpu::DepthStencilState depthStencilState = {}; + depthStencilState.depthWriteEnabled = IsDepthWriteEnabled; + + if (IsDepthTestEnabled) + { + depthStencilState.depthCompare = Convert(DepthFunc); + } + else + { + depthStencilState.depthCompare = wgpu::CompareFunction::Always; + } + + if (IsStencilTestEnabled) + { + wgpu::StencilFaceState fs; + + fs.compare = Convert(StencilCompareFunc); + fs.depthFailOp = Convert(StencilDepthFailOp); + fs.failOp = Convert(StencilFailOp); + fs.passOp = Convert(StencilPassOp); + + depthStencilState.stencilFront = fs; + depthStencilState.stencilBack = fs; + + depthStencilState.stencilWriteMask = StencilWriteMask; + depthStencilState.stencilReadMask = StencilReadMask; + } + else + { + wgpu::StencilFaceState fs; + + fs.depthFailOp = wgpu::StencilOperation::Keep; + fs.failOp = wgpu::StencilOperation::Keep; + fs.compare = wgpu::CompareFunction::Always; + fs.passOp = wgpu::StencilOperation::Keep; + + depthStencilState.stencilFront = fs; + depthStencilState.stencilBack = fs; + + depthStencilState.stencilWriteMask = 0xff; + depthStencilState.stencilReadMask = 0xff; + } + + if (renderPassPipelineState_->Key.DepthFormat != TextureFormatType::Unknown) + { + depthStencilState.format = ConvertFormat(renderPassPipelineState_->Key.DepthFormat); + desc.depthStencil = &depthStencilState; + } + + renderPipeline_ = device_.CreateRenderPipeline(&desc); + + return renderPipeline_ != nullptr; +} + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.PipelineStateWebGPU.h b/src/WebGPU/LLGI.PipelineStateWebGPU.h new file mode 100644 index 00000000..e0411774 --- /dev/null +++ b/src/WebGPU/LLGI.PipelineStateWebGPU.h @@ -0,0 +1,30 @@ +#pragma once + +#include "../LLGI.PipelineState.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ + +class PipelineStateWebGPU : public PipelineState +{ + std::array(ShaderStageType::Max)> shaders_; + + wgpu::Device device_; + + wgpu::RenderPipeline renderPipeline_; + wgpu::ComputePipeline computePipeline_; + +public: + PipelineStateWebGPU(wgpu::Device device); + ~PipelineStateWebGPU() override; + + void SetShader(ShaderStageType stage, Shader* shader) override; + + bool Compile() override; + + wgpu::RenderPipeline GetRenderPipeline() { return renderPipeline_; } + wgpu::ComputePipeline GetComputePipeline() { return computePipeline_; } +}; + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.PlatformWebGPU.cpp b/src/WebGPU/LLGI.PlatformWebGPU.cpp new file mode 100644 index 00000000..486b2e0f --- /dev/null +++ b/src/WebGPU/LLGI.PlatformWebGPU.cpp @@ -0,0 +1,381 @@ +#include "LLGI.PlatformWebGPU.h" +#include "LLGI.GraphicsWebGPU.h" +#include "LLGI.RenderPassWebGPU.h" +#include "LLGI.TextureWebGPU.h" + +#include +#include + +#if defined(_WIN32) +#ifndef NOMINMAX +#define NOMINMAX +#endif +#include +#endif + +namespace LLGI +{ + +namespace +{ +bool IsSurfaceTextureAcquired(wgpu::SurfaceGetCurrentTextureStatus status) +{ + return status == wgpu::SurfaceGetCurrentTextureStatus::SuccessOptimal || + status == wgpu::SurfaceGetCurrentTextureStatus::SuccessSuboptimal; +} + +wgpu::PresentMode SelectPresentMode(const wgpu::SurfaceCapabilities& capabilities, bool waitVSync) +{ + if (waitVSync) + { + return wgpu::PresentMode::Fifo; + } + + for (size_t i = 0; i < capabilities.presentModeCount; i++) + { + if (capabilities.presentModes[i] == wgpu::PresentMode::Immediate) + { + return wgpu::PresentMode::Immediate; + } + } + + for (size_t i = 0; i < capabilities.presentModeCount; i++) + { + if (capabilities.presentModes[i] == wgpu::PresentMode::Mailbox) + { + return wgpu::PresentMode::Mailbox; + } + } + + return wgpu::PresentMode::Fifo; +} +} // namespace + +PlatformWebGPU::PlatformWebGPU(wgpu::Device device) : device_(device) {} + +PlatformWebGPU::~PlatformWebGPU() { ResetCurrentScreen(); } + +void PlatformWebGPU::ResetCurrentScreen() +{ +#if !defined(__EMSCRIPTEN__) + if (surface_ != nullptr && surfaceTexture_.texture != nullptr && isPresentRequested_ && !hasPresentedCurrentSurface_) + { + surface_.Present(); + hasPresentedCurrentSurface_ = true; + } +#endif + + SafeRelease(currentScreenRenderPass_); + SafeRelease(currentScreenTexture_); + surfaceTexture_ = {}; + hasPresentedCurrentSurface_ = false; + isPresentRequested_ = false; +} + +bool PlatformWebGPU::ConfigureSurface(const Vec2I& windowSize) +{ + if (surface_ == nullptr || device_ == nullptr || windowSize.X <= 0 || windowSize.Y <= 0) + { + return false; + } +#if !defined(__EMSCRIPTEN__) + if (adapter_ == nullptr) + { + return false; + } +#endif + + wgpu::SurfaceCapabilities capabilities{}; + surface_.GetCapabilities(adapter_, &capabilities); + if (capabilities.formatCount == 0) + { + Log(LogType::Error, "WebGPU surface has no supported formats."); + return false; + } + + surfaceFormat_ = capabilities.formats[0]; + presentMode_ = SelectPresentMode(capabilities, waitVSync_); + + wgpu::SurfaceConfiguration config{}; + config.device = device_; + config.format = surfaceFormat_; + config.width = static_cast(windowSize.X); + config.height = static_cast(windowSize.Y); + config.presentMode = presentMode_; + config.usage = wgpu::TextureUsage::RenderAttachment | wgpu::TextureUsage::CopySrc | wgpu::TextureUsage::CopyDst; + surface_.Configure(&config); + + windowSize_ = windowSize; + return true; +} + +bool PlatformWebGPU::Initialize(Window* window, bool waitVSync) +{ + waitVSync_ = waitVSync; + window_ = window; +#if !defined(__EMSCRIPTEN__) + if (window_ == nullptr) + { + return false; + } +#endif + +#if defined(__EMSCRIPTEN__) + wgpu::InstanceDescriptor instanceDescriptor{}; + instance_ = wgpu::CreateInstance(&instanceDescriptor); + if (instance_ == nullptr) + { + Log(LogType::Error, "Failed to create browser WebGPU instance."); + return false; + } + + auto device = wgpu::Device::Acquire(emscripten_webgpu_get_device()); + if (device == nullptr) + { + Log(LogType::Error, "Failed to get preinitialized browser WebGPU device."); + return false; + } + + device_ = device; + if (window_ != nullptr) + { + windowSize_ = window_->GetWindowSize(); + + wgpu::EmscriptenSurfaceSourceCanvasHTMLSelector canvasSource{}; + canvasSource.selector = "#canvas"; + + wgpu::SurfaceDescriptor surfaceDescriptor{}; + surfaceDescriptor.nextInChain = &canvasSource; + surface_ = instance_.CreateSurface(&surfaceDescriptor); + if (surface_ == nullptr) + { + Log(LogType::Error, "Failed to create browser WebGPU canvas surface."); + return false; + } + + return ConfigureSurface(windowSize_); + } + return true; +#elif defined(_WIN32) + wgpu::InstanceDescriptor instanceDescriptor{}; + static constexpr auto timedWaitAny = wgpu::InstanceFeatureName::TimedWaitAny; + instanceDescriptor.requiredFeatureCount = 1; + instanceDescriptor.requiredFeatures = &timedWaitAny; + instance_ = wgpu::CreateInstance(&instanceDescriptor); + if (instance_ == nullptr) + { + Log(LogType::Error, "Failed to create WebGPU instance."); + return false; + } + + wgpu::SurfaceSourceWindowsHWND hwndSource{}; + hwndSource.hinstance = GetModuleHandleW(nullptr); + hwndSource.hwnd = window_->GetNativePtr(0); + + wgpu::SurfaceDescriptor surfaceDescriptor{}; + surfaceDescriptor.nextInChain = &hwndSource; + surface_ = instance_.CreateSurface(&surfaceDescriptor); + if (surface_ == nullptr) + { + Log(LogType::Error, "Failed to create WebGPU surface."); + return false; + } + + wgpu::RequestAdapterOptions adapterOptions{}; + adapterOptions.compatibleSurface = surface_; + adapterOptions.powerPreference = wgpu::PowerPreference::HighPerformance; + instance_.WaitAny( + instance_.RequestAdapter(&adapterOptions, + wgpu::CallbackMode::WaitAnyOnly, + [this](wgpu::RequestAdapterStatus status, wgpu::Adapter adapter, wgpu::StringView message) { + if (status != wgpu::RequestAdapterStatus::Success) + { + Log(LogType::Error, + std::string("Failed to request WebGPU adapter: ") + std::string(message.data, message.length)); + return; + } + adapter_ = adapter; + }), + std::numeric_limits::max()); + if (adapter_ == nullptr) + { + return false; + } + + wgpu::DeviceDescriptor deviceDescriptor{}; + std::vector requiredFeatures; + for (auto feature : {wgpu::FeatureName::Float32Filterable, wgpu::FeatureName::TextureFormatsTier2}) + { + if (adapter_.HasFeature(feature)) + { + requiredFeatures.push_back(feature); + } + } + deviceDescriptor.requiredFeatureCount = requiredFeatures.size(); + deviceDescriptor.requiredFeatures = requiredFeatures.data(); + deviceDescriptor.SetUncapturedErrorCallback([](const wgpu::Device&, wgpu::ErrorType, wgpu::StringView message) { + Log(LogType::Error, std::string("WebGPU validation error: ") + std::string(message.data, message.length)); + }); + + instance_.WaitAny( + adapter_.RequestDevice(&deviceDescriptor, + wgpu::CallbackMode::WaitAnyOnly, + [this](wgpu::RequestDeviceStatus status, wgpu::Device device, wgpu::StringView message) { + if (status != wgpu::RequestDeviceStatus::Success) + { + Log(LogType::Error, + std::string("Failed to request WebGPU device: ") + std::string(message.data, message.length)); + return; + } + device_ = device; + }), + std::numeric_limits::max()); + if (device_ == nullptr) + { + return false; + } + + return ConfigureSurface(window_->GetWindowSize()); +#else + Log(LogType::Error, "WebGPU window platform initialization is implemented for Windows only."); + return false; +#endif +} + +bool PlatformWebGPU::Initialize(wgpu::Device device, bool waitVSync) +{ + waitVSync_ = waitVSync; + device_ = device; + return device_ != nullptr; +} + +int PlatformWebGPU::GetCurrentFrameIndex() const { return 0; } + +int PlatformWebGPU::GetMaxFrameCount() const { return 1; } + +bool PlatformWebGPU::NewFrame() +{ + if (device_ == nullptr) + { + return false; + } + + if (window_ != nullptr) + { + if (!window_->OnNewFrame()) + { + return false; + } + + const auto windowSize = window_->GetWindowSize(); + if (windowSize != windowSize_) + { + SetWindowSize(windowSize); + } + } + + if (surface_ == nullptr) + { + return true; + } + + ResetCurrentScreen(); + return true; +} + +bool PlatformWebGPU::AcquireCurrentScreen() +{ + if (surface_ == nullptr) + { + return false; + } + + if (currentScreenRenderPass_ != nullptr) + { + return true; + } + + surface_.GetCurrentTexture(&surfaceTexture_); + if (!IsSurfaceTextureAcquired(surfaceTexture_.status) || surfaceTexture_.texture == nullptr) + { + if (surfaceTexture_.status == wgpu::SurfaceGetCurrentTextureStatus::Outdated || + surfaceTexture_.status == wgpu::SurfaceGetCurrentTextureStatus::Lost) + { + ConfigureSurface(windowSize_); + } + return false; + } + + TextureParameter textureParameter{}; + textureParameter.Usage = TextureUsageType::RenderTarget; + textureParameter.Format = ConvertFormat(surfaceFormat_); + textureParameter.Dimension = 2; + textureParameter.Size = Vec3I(windowSize_.X, windowSize_.Y, 1); + textureParameter.MipLevelCount = 1; + textureParameter.SampleCount = 1; + + currentScreenTexture_ = new TextureWebGPU(); + if (!currentScreenTexture_->InitializeFromSurfaceTexture(device_, surfaceTexture_.texture, textureParameter)) + { + ResetCurrentScreen(); + return false; + } + + Texture* textures[] = {currentScreenTexture_}; + currentScreenRenderPass_ = new RenderPassWebGPU(); + if (!currentScreenRenderPass_->Initialize(textures, 1, nullptr, nullptr, nullptr)) + { + ResetCurrentScreen(); + return false; + } + + return true; +} + +void PlatformWebGPU::Present() +{ + if (surface_ != nullptr && surfaceTexture_.texture != nullptr) + { + isPresentRequested_ = true; + } +} + +Graphics* PlatformWebGPU::CreateGraphics() +{ + if (device_ == nullptr) + { + return nullptr; + } + auto ret = new GraphicsWebGPU(device_, instance_); + ret->SetWindowSize(windowSize_); + return ret; +} + +void PlatformWebGPU::SetWindowSize(const Vec2I& windowSize) +{ + if (windowSize == windowSize_) + { + return; + } + ConfigureSurface(windowSize); +} + +RenderPass* PlatformWebGPU::GetCurrentScreen(const Color8& clearColor, bool isColorCleared, bool isDepthCleared) +{ + if (currentScreenRenderPass_ == nullptr) + { + AcquireCurrentScreen(); + } + + if (currentScreenRenderPass_ == nullptr) + { + return nullptr; + } + + currentScreenRenderPass_->SetClearColor(clearColor); + currentScreenRenderPass_->SetIsColorCleared(isColorCleared); + currentScreenRenderPass_->SetIsDepthCleared(isDepthCleared); + return currentScreenRenderPass_; +} + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.PlatformWebGPU.h b/src/WebGPU/LLGI.PlatformWebGPU.h new file mode 100644 index 00000000..fbad8d1b --- /dev/null +++ b/src/WebGPU/LLGI.PlatformWebGPU.h @@ -0,0 +1,52 @@ +#pragma once + +#include "../LLGI.Platform.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ + +class RenderPassWebGPU; +class TextureWebGPU; + +class PlatformWebGPU : public Platform +{ +private: + Window* window_ = nullptr; + Vec2I windowSize_; + wgpu::Instance instance_; + wgpu::Adapter adapter_; + wgpu::Device device_; + wgpu::Surface surface_; + wgpu::TextureFormat surfaceFormat_ = wgpu::TextureFormat::Undefined; + wgpu::PresentMode presentMode_ = wgpu::PresentMode::Fifo; + wgpu::SurfaceTexture surfaceTexture_; + TextureWebGPU* currentScreenTexture_ = nullptr; + RenderPassWebGPU* currentScreenRenderPass_ = nullptr; + bool hasPresentedCurrentSurface_ = false; + bool isPresentRequested_ = false; + + bool ConfigureSurface(const Vec2I& windowSize); + bool AcquireCurrentScreen(); + void ResetCurrentScreen(); + +public: + PlatformWebGPU() = default; + explicit PlatformWebGPU(wgpu::Device device); + ~PlatformWebGPU() override; + + bool Initialize(Window* window, bool waitVSync); + bool Initialize(wgpu::Device device, bool waitVSync); + + int GetCurrentFrameIndex() const override; + int GetMaxFrameCount() const override; + + bool NewFrame() override; + void Present() override; + Graphics* CreateGraphics() override; + DeviceType GetDeviceType() const override { return DeviceType::WebGPU; } + void SetWindowSize(const Vec2I& windowSize) override; + RenderPass* GetCurrentScreen(const Color8& clearColor, bool isColorCleared, bool isDepthCleared) override; +}; + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.RenderPassPipelineStateWebGPU.cpp b/src/WebGPU/LLGI.RenderPassPipelineStateWebGPU.cpp new file mode 100644 index 00000000..fc02a044 --- /dev/null +++ b/src/WebGPU/LLGI.RenderPassPipelineStateWebGPU.cpp @@ -0,0 +1,19 @@ +#include "LLGI.RenderPassPipelineStateWebGPU.h" + +namespace LLGI +{ + +void RenderPassPipelineStateWebGPU::SetKey(const RenderPassPipelineStateKey& key) +{ + Key = key; + pixelFormats_.resize(key.RenderTargetFormats.size()); + + for (size_t i = 0; i < pixelFormats_.size(); i++) + { + pixelFormats_.at(i) = ConvertFormat(key.RenderTargetFormats.at(i)); + } + + depthStencilFormat_ = ConvertFormat(key.DepthFormat); +} + +} // namespace LLGI \ No newline at end of file diff --git a/src/WebGPU/LLGI.RenderPassPipelineStateWebGPU.h b/src/WebGPU/LLGI.RenderPassPipelineStateWebGPU.h new file mode 100644 index 00000000..ed962efe --- /dev/null +++ b/src/WebGPU/LLGI.RenderPassPipelineStateWebGPU.h @@ -0,0 +1,21 @@ +#pragma once + +#include "../LLGI.Graphics.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ +class RenderPassPipelineStateWebGPU : public RenderPassPipelineState +{ +private: + FixedSizeVector pixelFormats_; + wgpu::TextureFormat depthStencilFormat_ = wgpu::TextureFormat::Undefined; + +public: + void SetKey(const RenderPassPipelineStateKey& key); + + const FixedSizeVector& GetPixelFormats() const { return pixelFormats_; } + const wgpu::TextureFormat& GetDepthStencilFormat() const { return depthStencilFormat_; } +}; + +} // namespace LLGI \ No newline at end of file diff --git a/src/WebGPU/LLGI.RenderPassWebGPU.cpp b/src/WebGPU/LLGI.RenderPassWebGPU.cpp new file mode 100644 index 00000000..08b73e7a --- /dev/null +++ b/src/WebGPU/LLGI.RenderPassWebGPU.cpp @@ -0,0 +1,139 @@ +#include "LLGI.RenderPassWebGPU.h" +#include "LLGI.TextureWebGPU.h" + +namespace LLGI +{ + +void RenderPassWebGPU::RefreshDescriptor() +{ + for (int i = 0; i < descriptor_.colorAttachmentCount; i++) + { + if (GetIsColorCleared()) + { + colorAttachments_[i].loadOp = wgpu::LoadOp::Clear; + colorAttachments_[i].storeOp = wgpu::StoreOp::Store; + colorAttachments_[i].clearValue = { + GetClearColor().R / 255.0, GetClearColor().G / 255.0, GetClearColor().B / 255.0, GetClearColor().A / 255.0}; + } + else + { + colorAttachments_[i].loadOp = wgpu::LoadOp::Load; + colorAttachments_[i].storeOp = wgpu::StoreOp::Store; + colorAttachments_[i].clearValue = {0, 0, 0, 1}; + } + } + + if (descriptor_.depthStencilAttachment != nullptr) + { + if (GetIsDepthCleared()) + { + depthStencilAttachiment_.depthLoadOp = wgpu::LoadOp::Clear; + depthStencilAttachiment_.depthStoreOp = wgpu::StoreOp::Store; + depthStencilAttachiment_.depthClearValue = 1.0f; + } + else + { + depthStencilAttachiment_.depthLoadOp = wgpu::LoadOp::Load; + depthStencilAttachiment_.depthStoreOp = wgpu::StoreOp::Store; + depthStencilAttachiment_.depthClearValue = 1.0f; + } + } +} + +bool RenderPassWebGPU::Initialize( + Texture** textures, int textureCount, Texture* depthTexture, Texture* resolvedRenderTexture, Texture* resolvedDepthTexture) +{ + if (!assignRenderTextures(textures, textureCount)) + { + return false; + } + + if (!assignDepthTexture(depthTexture)) + { + return false; + } + + if (!assignResolvedRenderTexture(resolvedRenderTexture)) + { + return false; + } + + if (!assignResolvedDepthTexture(resolvedDepthTexture)) + { + return false; + } + + if (!getSize(screenSize_, (const Texture**)textures, textureCount, depthTexture)) + { + return false; + } + + std::array texturesImpl; + texturesImpl.fill(nullptr); + TextureWebGPU* depthTextureImpl = nullptr; + + for (int32_t i = 0; i < textureCount; i++) + { + if (textures[i] == nullptr) + continue; + + texturesImpl.at(i) = reinterpret_cast(textures[i]); + } + + if (depthTexture != nullptr) + { + depthTextureImpl = reinterpret_cast(depthTexture); + } + + TextureWebGPU* resolvedTextureImpl = nullptr; + TextureWebGPU* resolvedDepthTextureImpl = nullptr; + + if (resolvedRenderTexture != nullptr) + { + resolvedTextureImpl = reinterpret_cast(resolvedRenderTexture); + } + + if (resolvedDepthTexture != nullptr) + { + resolvedDepthTextureImpl = reinterpret_cast(resolvedDepthTexture); + } + + descriptor_.colorAttachmentCount = textureCount; + descriptor_.colorAttachments = colorAttachments_.data(); + + for (int i = 0; i < textureCount; i++) + { + colorAttachments_[i].view = texturesImpl[i]->GetTextureView(); + + if (resolvedTextureImpl != nullptr) + { + colorAttachments_[i].resolveTarget = resolvedTextureImpl->GetTextureView(); + colorAttachments_[i].storeOp = wgpu::StoreOp::Store; + } + } + + if (depthTexture != nullptr) + { + depthStencilAttachiment_.view = depthTextureImpl->GetTextureView(); + + if (depthTextureImpl->GetFormat() == TextureFormatType::D24S8 || depthTextureImpl->GetFormat() == TextureFormatType::D32S8) + { + depthStencilAttachiment_.stencilLoadOp = GetIsDepthCleared() ? wgpu::LoadOp::Clear : wgpu::LoadOp::Load; + depthStencilAttachiment_.stencilStoreOp = wgpu::StoreOp::Store; + depthStencilAttachiment_.stencilClearValue = 0; + } + + if (resolvedDepthTextureImpl != nullptr) + { + // ? + } + + descriptor_.depthStencilAttachment = &depthStencilAttachiment_; + } + + RefreshDescriptor(); + + return true; +} + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.RenderPassWebGPU.h b/src/WebGPU/LLGI.RenderPassWebGPU.h new file mode 100644 index 00000000..23ee7d46 --- /dev/null +++ b/src/WebGPU/LLGI.RenderPassWebGPU.h @@ -0,0 +1,25 @@ +#pragma once + +#include "../LLGI.Graphics.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ +class TextureWebGPU; + +class RenderPassWebGPU : public RenderPass +{ + wgpu::RenderPassDescriptor descriptor_; + std::array colorAttachments_; + wgpu::RenderPassDepthStencilAttachment depthStencilAttachiment_; + +public: + bool + Initialize(Texture** textures, int textureCount, Texture* depthTexture, Texture* resolvedRenderTexture, Texture* resolvedDepthTexture); + + void RefreshDescriptor(); + + const wgpu::RenderPassDescriptor& GetDescriptor() const { return descriptor_; } +}; + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.ShaderWebGPU.cpp b/src/WebGPU/LLGI.ShaderWebGPU.cpp new file mode 100644 index 00000000..dc01a7d7 --- /dev/null +++ b/src/WebGPU/LLGI.ShaderWebGPU.cpp @@ -0,0 +1,66 @@ +#include "LLGI.ShaderWebGPU.h" +#include +#include + +namespace LLGI +{ +ShaderWebGPU::ShaderWebGPU() {} + +ShaderWebGPU::~ShaderWebGPU() {} + +bool ShaderWebGPU::Initialize(wgpu::Device& device, DataStructure* data, int32_t count) +{ + static const char wgslHeader[] = {'w', 'g', 's', 'l', 'c', 'o', 'd', 'e'}; + + if (data == nullptr || count == 0) + { + return false; + } + + wgpu::ShaderModuleDescriptor desc = {}; + + if (data[0].Data == nullptr || data[0].Size <= 0) + { + return false; + } + + const auto* bytes = static_cast(data[0].Data); + const bool hasWGSLHeader = data[0].Size >= static_cast(sizeof(wgslHeader)) && + memcmp(bytes, wgslHeader, sizeof(wgslHeader)) == 0; + const bool hasSPIRVMagic = data[0].Size >= 4 && bytes[0] == 0x03 && bytes[1] == 0x02 && bytes[2] == 0x23 && bytes[3] == 0x07; + + wgpu::ShaderSourceSPIRV sprivDesc = {}; + wgpu::ShaderSourceWGSL wgslDesc = {}; + std::string wgslCode; + + if (!hasSPIRVMagic) + { + const auto codeOffset = hasWGSLHeader ? sizeof(wgslHeader) : 0; + if (data[0].Size <= static_cast(codeOffset)) + { + return false; + } + + wgslCode.assign(reinterpret_cast(bytes + codeOffset), static_cast(data[0].Size) - codeOffset); + while (!wgslCode.empty() && wgslCode.back() == '\0') + { + wgslCode.pop_back(); + } + wgslDesc.code = wgpu::StringView(wgslCode.data(), wgslCode.size()); + desc.nextInChain = reinterpret_cast(&wgslDesc); + } + else + { + sprivDesc.codeSize = data[0].Size / sizeof(uint32_t); + sprivDesc.code = reinterpret_cast(data[0].Data); + desc.nextInChain = reinterpret_cast(&sprivDesc); + } + + shaderModule_ = device.CreateShaderModule(&desc); + + return shaderModule_ != nullptr; +} + +wgpu::ShaderModule& ShaderWebGPU::GetShaderModule() { return shaderModule_; } + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.ShaderWebGPU.h b/src/WebGPU/LLGI.ShaderWebGPU.h new file mode 100644 index 00000000..0cd74907 --- /dev/null +++ b/src/WebGPU/LLGI.ShaderWebGPU.h @@ -0,0 +1,23 @@ +#pragma once + +#include "LLGI.BaseWebGPU.h" +#include "../LLGI.Shader.h" + +namespace LLGI +{ + +class ShaderWebGPU : public Shader +{ +private: + wgpu::ShaderModule shaderModule_; + +public: + ShaderWebGPU(); + ~ShaderWebGPU() override; + + bool Initialize(wgpu::Device& device, DataStructure* data, int32_t count); + + wgpu::ShaderModule& GetShaderModule(); +}; + +} \ No newline at end of file diff --git a/src/WebGPU/LLGI.TextureWebGPU.cpp b/src/WebGPU/LLGI.TextureWebGPU.cpp new file mode 100644 index 00000000..eb66a3a0 --- /dev/null +++ b/src/WebGPU/LLGI.TextureWebGPU.cpp @@ -0,0 +1,284 @@ +#include "LLGI.TextureWebGPU.h" + +#include +#include +#include + +#if defined(__EMSCRIPTEN__) +#include +#endif + +namespace LLGI +{ + +namespace +{ +uint32_t AlignTo(uint32_t value, uint32_t alignment) +{ + return (value + alignment - 1) / alignment * alignment; +} +} // namespace + +bool TextureWebGPU::Initialize(wgpu::Device& device, const TextureParameter& parameter, wgpu::Instance instance) +{ + device_ = device; + instance_ = instance; + parameter_ = parameter; + + const auto getDimension = [](int dimension) + { + if (dimension == 1) + return wgpu::TextureDimension::e1D; + + if (dimension == 2) + return wgpu::TextureDimension::e2D; + + if (dimension == 3) + return wgpu::TextureDimension::e3D; + + throw "Not implemented"; + }; + + const auto getViewDimension = [](int dimension) + { + if (dimension == 1) + return wgpu::TextureViewDimension::e1D; + + if (dimension == 2) + return wgpu::TextureViewDimension::e2D; + + if (dimension == 3) + return wgpu::TextureViewDimension::e3D; + + throw "Not implemented"; + }; + + { + wgpu::TextureDescriptor texDesc{}; + + texDesc.usage = wgpu::TextureUsage::TextureBinding | wgpu::TextureUsage::CopyDst | wgpu::TextureUsage::CopySrc; + if ((parameter.Usage & TextureUsageType::RenderTarget) != TextureUsageType::NoneFlag) + { + texDesc.usage |= wgpu::TextureUsage::RenderAttachment; + } + + if (BitwiseContains(parameter.Usage, TextureUsageType::Storage)) + { + texDesc.usage |= wgpu::TextureUsage::StorageBinding; + } + + if ((parameter.Usage & TextureUsageType::External) != TextureUsageType::NoneFlag) + { + throw "Not implemented"; + // texDesc.usage |= dawn::platform::kPresentTextureUsage; + } + + bool isArray = false; + if ((parameter.Usage & TextureUsageType::Array) != TextureUsageType::NoneFlag) + { + isArray = true; + } + + texDesc.dimension = getDimension(parameter.Dimension); + texDesc.format = ConvertFormat(parameter.Format); + texDesc.mipLevelCount = parameter.MipLevelCount; + texDesc.sampleCount = parameter.SampleCount; + texDesc.size.width = parameter.Size.X; + texDesc.size.height = parameter.Size.Y; + texDesc.size.depthOrArrayLayers = parameter.Size.Z; + + texture_ = device.CreateTexture(&texDesc); + if (texture_ == nullptr) + { + return false; + } + + wgpu::TextureViewDescriptor texViewDesc{}; + texViewDesc.format = texDesc.format; + texViewDesc.dimension = isArray && parameter.Dimension == 2 ? wgpu::TextureViewDimension::e2DArray : getViewDimension(parameter.Dimension); + texViewDesc.baseMipLevel = 0; + texViewDesc.mipLevelCount = texDesc.mipLevelCount; + texViewDesc.baseArrayLayer = 0; + texViewDesc.arrayLayerCount = isArray ? parameter.Size.Z : 1; + texViewDesc.aspect = wgpu::TextureAspect::All; + + textureView_ = texture_.CreateView(&texViewDesc); + } + + format_ = parameter.Format; + usage_ = parameter.Usage; + samplingCount_ = parameter.SampleCount; + mipmapCount_ = parameter.MipLevelCount; + type_ = TextureType::Color; + if (IsDepthFormat(parameter.Format)) + { + type_ = TextureType::Depth; + } + else if (BitwiseContains(parameter.Usage, TextureUsageType::RenderTarget)) + { + type_ = TextureType::Render; + } + + return texture_ != nullptr && textureView_ != nullptr; +} + +bool TextureWebGPU::InitializeFromSurfaceTexture(wgpu::Device& device, wgpu::Texture texture, const TextureParameter& parameter) +{ + device_ = device; + parameter_ = parameter; + texture_ = texture; + if (texture_ == nullptr) + { + return false; + } + + wgpu::TextureViewDescriptor texViewDesc{}; + texViewDesc.format = ConvertFormat(parameter.Format); + texViewDesc.dimension = wgpu::TextureViewDimension::e2D; + texViewDesc.baseMipLevel = 0; + texViewDesc.mipLevelCount = 1; + texViewDesc.baseArrayLayer = 0; + texViewDesc.arrayLayerCount = 1; + texViewDesc.aspect = wgpu::TextureAspect::All; + textureView_ = texture_.CreateView(&texViewDesc); + + format_ = parameter.Format; + usage_ = parameter.Usage; + samplingCount_ = parameter.SampleCount; + mipmapCount_ = parameter.MipLevelCount; + type_ = TextureType::Screen; + + return textureView_ != nullptr; +} + +void* TextureWebGPU::Lock() +{ + auto cpuMemorySize = GetTextureMemorySize(format_, parameter_.Size); + temp_buffer_.resize(cpuMemorySize); + return temp_buffer_.data(); +} + +void TextureWebGPU::Unlock() +{ + wgpu::TexelCopyTextureInfo imageCopyTexture{}; + imageCopyTexture.texture = texture_; + imageCopyTexture.aspect = wgpu::TextureAspect::All; + + wgpu::TexelCopyBufferLayout textureDataLayout; + textureDataLayout.bytesPerRow = GetTextureRowPitch(format_, parameter_.Size); + wgpu::Extent3D extent; + extent.width = parameter_.Size.X; + extent.height = parameter_.Size.Y; + extent.depthOrArrayLayers = parameter_.Size.Z; + device_.GetQueue().WriteTexture(&imageCopyTexture, temp_buffer_.data(), temp_buffer_.size(), &textureDataLayout, &extent); +} + +bool TextureWebGPU::GetData(std::vector& data) +{ + const auto bytesPerRowUnaligned = static_cast(GetTextureRowPitch(format_, parameter_.Size)); + const auto bytesPerRow = AlignTo(bytesPerRowUnaligned, 256); + const auto height = static_cast(parameter_.Size.Y); + const auto depth = static_cast(parameter_.Size.Z); + const auto bufferSize = static_cast(bytesPerRow) * height * depth; + + wgpu::BufferDescriptor bufferDesc{}; + bufferDesc.size = bufferSize; + bufferDesc.usage = wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::MapRead; + auto readbackBuffer = device_.CreateBuffer(&bufferDesc); + + wgpu::CommandEncoderDescriptor encoderDesc{}; + auto encoder = device_.CreateCommandEncoder(&encoderDesc); + + wgpu::TexelCopyTextureInfo src{}; + src.texture = texture_; + src.aspect = wgpu::TextureAspect::All; + + wgpu::TexelCopyBufferInfo dst{}; + dst.buffer = readbackBuffer; + dst.layout.bytesPerRow = bytesPerRow; + dst.layout.rowsPerImage = height; + + wgpu::Extent3D extent{}; + extent.width = static_cast(parameter_.Size.X); + extent.height = height; + extent.depthOrArrayLayers = depth; + encoder.CopyTextureToBuffer(&src, &dst, &extent); + + auto commandBuffer = encoder.Finish(); + device_.GetQueue().Submit(1, &commandBuffer); + + bool completed = false; + bool succeeded = false; + auto future = readbackBuffer.MapAsync(wgpu::MapMode::Read, + 0, + bufferSize, +#if defined(__EMSCRIPTEN__) + wgpu::CallbackMode::AllowSpontaneous, +#else + instance_ != nullptr ? wgpu::CallbackMode::WaitAnyOnly : wgpu::CallbackMode::AllowProcessEvents, +#endif + [&completed, &succeeded](wgpu::MapAsyncStatus status, wgpu::StringView) { + succeeded = status == wgpu::MapAsyncStatus::Success; + completed = true; + }); + + if (instance_ != nullptr) + { + instance_.WaitAny(future, 5ULL * 1000ULL * 1000ULL * 1000ULL); + } + else + { +#if defined(__EMSCRIPTEN__) + const double waitStart = emscripten_get_now(); + while (!completed) + { + emscripten_sleep(1); + if (emscripten_get_now() - waitStart > 5000.0) + { + break; + } + } +#else + const auto waitStart = std::chrono::steady_clock::now(); + while (!completed) + { + device_.Tick(); + if (std::chrono::steady_clock::now() - waitStart > std::chrono::seconds(5)) + { + break; + } + std::this_thread::sleep_for(std::chrono::milliseconds(1)); + } +#endif + } + + if (!succeeded) + { + return false; + } + + data.resize(static_cast(bytesPerRowUnaligned) * height * depth); + const auto mapped = static_cast(readbackBuffer.GetConstMappedRange(0, bufferSize)); + for (uint32_t z = 0; z < depth; z++) + { + for (uint32_t y = 0; y < height; y++) + { + memcpy(data.data() + (static_cast(z) * height + y) * bytesPerRowUnaligned, + mapped + (static_cast(z) * height + y) * bytesPerRow, + bytesPerRowUnaligned); + } + } + readbackBuffer.Unmap(); + return true; +} + +Vec2I TextureWebGPU::GetSizeAs2D() const { return Vec2I(parameter_.Size.X, parameter_.Size.Y); } + +bool TextureWebGPU::IsRenderTexture() const +{ + return type_ == TextureType::Render || type_ == TextureType::Screen; +} + +bool TextureWebGPU::IsDepthTexture() const { return type_ == TextureType::Depth; } + +} // namespace LLGI diff --git a/src/WebGPU/LLGI.TextureWebGPU.h b/src/WebGPU/LLGI.TextureWebGPU.h new file mode 100644 index 00000000..76354bb9 --- /dev/null +++ b/src/WebGPU/LLGI.TextureWebGPU.h @@ -0,0 +1,37 @@ +#pragma once + +#include "../LLGI.Graphics.h" +#include "../LLGI.Texture.h" +#include "LLGI.BaseWebGPU.h" + +namespace LLGI +{ + +class TextureWebGPU : public Texture +{ + wgpu::Device device_; + wgpu::Instance instance_; + + wgpu::Texture texture_; + wgpu::TextureView textureView_; + TextureParameter parameter_; + std::vector temp_buffer_; + +public: + bool Initialize(wgpu::Device& device, const TextureParameter& parameter, wgpu::Instance instance = nullptr); + bool InitializeFromSurfaceTexture(wgpu::Device& device, wgpu::Texture texture, const TextureParameter& parameter); + void* Lock() override; + void Unlock() override; + bool GetData(std::vector& data) override; + Vec2I GetSizeAs2D() const override; + bool IsRenderTexture() const override; + bool IsDepthTexture() const override; + + const TextureParameter& GetParameter() const { return parameter_; } + + wgpu::Texture GetTexture() const { return texture_; } + + wgpu::TextureView GetTextureView() const { return textureView_; } +}; + +} // namespace LLGI diff --git a/src_test/CMakeLists.txt b/src_test/CMakeLists.txt index a256422b..5d47f5ce 100644 --- a/src_test/CMakeLists.txt +++ b/src_test/CMakeLists.txt @@ -1,7 +1,16 @@ -file(GLOB files *.h *.cpp) +if(BUILD_WEBGPU_BROWSER_TEST) + set(files main.cpp TestHelper.cpp TestHelper.h test.h test_webgpu_browser.cpp) +else() + file(GLOB files *.h *.cpp) + list(FILTER files EXCLUDE REGEX "test_webgpu_browser\\.cpp$") +endif() add_executable(LLGI_Test ${files}) +if(BUILD_WEBGPU_BROWSER_TEST AND NOT EMSCRIPTEN) + message(FATAL_ERROR "BUILD_WEBGPU_BROWSER_TEST requires an Emscripten CMake toolchain.") +endif() + if(APPLE) find_library(COCOA_LIBRARY Cocoa) @@ -29,7 +38,17 @@ if(BUILD_VULKAN_COMPILER AND USE_THIRDPARTY_DIRECTORY) add_dependencies(LLGI_Test EP_glslang EP_SPIRV-Cross) endif() -if(MSVC) +if(BUILD_WEBGPU_BROWSER_TEST) + set_target_properties(LLGI_Test PROPERTIES SUFFIX ".html") + target_link_options( + LLGI_Test + PRIVATE + "--pre-js=${CMAKE_CURRENT_SOURCE_DIR}/browser/pre_webgpu_test.js" + "--preload-file=${CMAKE_CURRENT_SOURCE_DIR}/Shaders/WebGPU@/Shaders/WebGPU" + "-sALLOW_MEMORY_GROWTH=1" + "-sASYNCIFY=1" + "-sASSERTIONS=1") +elseif(MSVC) target_link_libraries(LLGI_Test PRIVATE) elseif(APPLE) target_link_libraries(LLGI_Test PRIVATE) @@ -39,13 +58,27 @@ else() X11-xcb) endif() -file(COPY ${CMAKE_CURRENT_SOURCE_DIR}/Shaders - DESTINATION ${CMAKE_CURRENT_BINARY_DIR}/) +if(NOT EMSCRIPTEN) + file(COPY ${CMAKE_CURRENT_SOURCE_DIR}/Shaders + DESTINATION ${CMAKE_CURRENT_BINARY_DIR}/) +endif() clang_format(LLGI_Test) -if(MSVC) +if(EMSCRIPTEN) +elseif(MSVC) target_compile_options(LLGI_Test PRIVATE /W4 /WX /wd4100) else() target_compile_options(LLGI_Test PRIVATE -Wall -Werror) endif() + +if(BUILD_WEBGPU_BROWSER_TEST AND EMSCRIPTEN) + find_program(NODE_EXE node) + if(NODE_EXE) + add_test( + NAME LLGI_WebGPU_Browser + COMMAND + ${NODE_EXE} ${CMAKE_CURRENT_SOURCE_DIR}/browser/run_webgpu_browser_test.mjs + $ --filter=WebGPUBrowser.*) + endif() +endif() diff --git a/src_test/Shaders/WebGPU/basic.comp b/src_test/Shaders/WebGPU/basic.comp new file mode 100644 index 00000000..413bfbcd --- /dev/null +++ b/src_test/Shaders/WebGPU/basic.comp @@ -0,0 +1,42 @@ +diagnostic(off, derivative_uniformity); + +struct CS_OUTPUT { + value : f32, +} + +struct S { + m : array, +} + +@group(2) @binding(1) var write_1 : S; + +struct CS_INPUT { + value1 : f32, + value2 : f32, +} + +struct read_2 { + m_1 : array, +} + +@group(2) @binding(0) var read_1 : read_2; + +struct CB { + offset : f32, +} + +@group(0) @binding(0) var v : CB; + +@compute @workgroup_size(1u, 1u, 1u) +fn main(@builtin(global_invocation_id) dtid : vec3) { + var dtid_1 : vec3; + var param : vec3; + dtid_1 = dtid; + param = dtid_1; + v_1(&(param)); +} + +fn v_1(dtid : ptr>) { + let v_2 = (*(dtid)).x; + write_1.m[v_2].value = ((read_1.m_1[(*(dtid)).x].value1 * read_1.m_1[(*(dtid)).x].value2) + v.offset); +} diff --git a/src_test/Shaders/WebGPU/instancing.vert b/src_test/Shaders/WebGPU/instancing.vert new file mode 100644 index 00000000..0b88b11b --- /dev/null +++ b/src_test/Shaders/WebGPU/instancing.vert @@ -0,0 +1,62 @@ +diagnostic(off, derivative_uniformity); + +struct CB { + offsets : array, 10u>, +} + +@group(0) @binding(0) var v : CB; + +var v_1 : vec4; + +var v_2 : vec4; + +struct VS_INPUT { + g_position : vec3, + g_uv : vec2, + g_color : vec4, + InstanceId : u32, +} + +struct VS_OUTPUT { + g_position : vec4, + g_color : vec4, +} + +fn main_inner(v_3 : vec3, v_4 : vec2, v_5 : vec4, v_6 : u32) { + var input : VS_INPUT; + var flattenTemp : VS_OUTPUT; + var param : VS_INPUT; + input.g_position = v_3; + input.g_uv = v_4; + input.g_color = v_5; + input.InstanceId = v_6; + param = input; + flattenTemp = v_7(&(param)); + v_1 = flattenTemp.g_position; + v_2 = flattenTemp.g_color; +} + +fn v_7(input : ptr) -> VS_OUTPUT { + var output : VS_OUTPUT; + let v_8 = (*(input)).g_position; + output.g_position = vec4(v_8.x, v_8.y, v_8.z, 1.0f); + let v_9 = v.offsets[(*(input)).InstanceId].x; + output.g_position.x = (output.g_position.x + v_9); + let v_10 = v.offsets[(*(input)).InstanceId].y; + output.g_position.y = (output.g_position.y + v_10); + output.g_color = (*(input)).g_color; + return output; +} + +struct tint_symbol { + @builtin(position) + m : vec4, + @location(0u) + m_1 : vec4, +} + +@vertex +fn main(@location(0u) v_11 : vec3, @location(1u) v_12 : vec2, @location(2u) v_13 : vec4, @builtin(instance_index) v_14 : u32) -> tint_symbol { + main_inner(v_11, v_12, v_13, v_14); + return tint_symbol(v_1, v_2); +} diff --git a/src_test/Shaders/WebGPU/readwrite.comp b/src_test/Shaders/WebGPU/readwrite.comp new file mode 100644 index 00000000..c1c3839b --- /dev/null +++ b/src_test/Shaders/WebGPU/readwrite.comp @@ -0,0 +1,42 @@ +diagnostic(off, derivative_uniformity); + +struct CS_OUTPUT { + value : f32, +} + +struct S { + m : array, +} + +@group(2) @binding(1) var write_1 : S; + +struct CS_INPUT { + value1 : f32, + value2 : f32, +} + +struct read_2 { + m_1 : array, +} + +@group(2) @binding(0) var read_1 : read_2; + +struct CB { + offset : f32, +} + +@group(0) @binding(0) var v : CB; + +@compute @workgroup_size(1u, 1u, 1u) +fn main(@builtin(global_invocation_id) dtid : vec3) { + var dtid_1 : vec3; + var param : vec3; + dtid_1 = dtid; + param = dtid_1; + v_1(&(param)); +} + +fn v_1(dtid : ptr>) { + let v_2 = (*(dtid)).x; + write_1.m[v_2].value = ((read_1.m_1[(*(dtid)).x].value1 * read_1.m_1[(*(dtid)).x].value2) + v.offset); +} diff --git a/src_test/Shaders/WebGPU/readwrite_texture.comp b/src_test/Shaders/WebGPU/readwrite_texture.comp new file mode 100644 index 00000000..9be01f55 --- /dev/null +++ b/src_test/Shaders/WebGPU/readwrite_texture.comp @@ -0,0 +1,26 @@ +diagnostic(off, derivative_uniformity); + +@group(1) @binding(1) var read1 : texture_storage_2d; + +@group(1) @binding(2) var read2 : texture_2d; + +@group(2) @binding(2) var read2_sampler : sampler; + +@group(1) @binding(0) var v : texture_storage_2d; + +@compute @workgroup_size(1u, 1u, 1u) +fn main(@builtin(global_invocation_id) dtid : vec3) { + var dtid_1 : vec3; + var param : vec3; + dtid_1 = dtid; + param = dtid_1; + v_1(&(param)); +} + +fn v_1(dtid : ptr>) { + var index : vec2; + var storeTemp : vec4; + index = (*(dtid)).xy; + storeTemp = (textureLoad(read1, index) + textureSampleLevel(read2, read2_sampler, vec2(), 0.0f)); + textureStore(v, index, storeTemp); +} diff --git a/src_test/Shaders/WebGPU/simple_compute_rectangle.frag b/src_test/Shaders/WebGPU/simple_compute_rectangle.frag new file mode 100644 index 00000000..e10eb03f --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_compute_rectangle.frag @@ -0,0 +1,46 @@ +diagnostic(off, derivative_uniformity); + +struct compute { + m : array, +} + +@group(1) @binding(0) var compute_1 : compute; + +var v : vec4; + +struct CB { + offset : vec4, +} + +@group(0) @binding(1) var v_1 : CB; + +struct PS_INPUT { + Position : vec4, + UV : vec2, + Color : vec4, +} + +fn main_inner(v_2 : vec4, v_3 : vec2, v_4 : vec4) { + var input : PS_INPUT; + var param : PS_INPUT; + input.Position = v_2; + input.UV = v_3; + input.Color = v_4; + param = input; + v = v_5(&(param)); +} + +fn v_5(input : ptr) -> vec4 { + var c : vec4; + let v_6 = (*(input)).Color; + let v_7 = compute_1.m[0i]; + c = (v_6 + vec4(v_7, v_7, v_7, v_7)); + c.w = 1.0f; + return c; +} + +@fragment +fn main(@builtin(position) v_8 : vec4, @location(0u) v_9 : vec2, @location(1u) v_10 : vec4) -> @location(0u) vec4 { + main_inner(v_8, v_9, v_10); + return v; +} diff --git a/src_test/Shaders/WebGPU/simple_compute_rectangle.vert b/src_test/Shaders/WebGPU/simple_compute_rectangle.vert new file mode 100644 index 00000000..7229e532 --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_compute_rectangle.vert @@ -0,0 +1,64 @@ +diagnostic(off, derivative_uniformity); + +struct compute { + m : array, +} + +@group(1) @binding(0) var compute_1 : compute; + +var v : vec4; + +var v_1 : vec2; + +var v_2 : vec4; + +struct VS_INPUT { + Position : vec3, + UV : vec2, + Color : vec4, +} + +struct VS_OUTPUT { + Position : vec4, + UV : vec2, + Color : vec4, +} + +fn main_inner(v_3 : vec3, v_4 : vec2, v_5 : vec4) { + var input : VS_INPUT; + var flattenTemp : VS_OUTPUT; + var param : VS_INPUT; + input.Position = v_3; + input.UV = v_4; + input.Color = v_5; + param = input; + flattenTemp = v_6(&(param)); + v = flattenTemp.Position; + v_1 = flattenTemp.UV; + v_2 = flattenTemp.Color; +} + +fn v_6(input : ptr) -> VS_OUTPUT { + var output : VS_OUTPUT; + let v_7 = (*(input)).Position; + let v_8 = compute_1.m[0i]; + output.Position = (vec4(v_7.x, v_7.y, v_7.z, 1.0f) + vec4(v_8, v_8, v_8, v_8)); + output.UV = (*(input)).UV; + output.Color = (*(input)).Color; + return output; +} + +struct tint_symbol { + @builtin(position) + m_1 : vec4, + @location(0u) + m_2 : vec2, + @location(1u) + m_3 : vec4, +} + +@vertex +fn main(@location(0u) v_9 : vec3, @location(1u) v_10 : vec2, @location(2u) v_11 : vec4) -> tint_symbol { + main_inner(v_9, v_10, v_11); + return tint_symbol(v, v_1, v_2); +} diff --git a/src_test/Shaders/WebGPU/simple_constant_rectangle.frag b/src_test/Shaders/WebGPU/simple_constant_rectangle.frag new file mode 100644 index 00000000..17bbaf8a --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_constant_rectangle.frag @@ -0,0 +1,38 @@ +diagnostic(off, derivative_uniformity); + +struct CB { + offset : vec4, +} + +@group(0) @binding(1) var v : CB; + +var v_1 : vec4; + +struct PS_INPUT { + Position : vec4, + UV : vec2, + Color : vec4, +} + +fn main_inner(v_2 : vec4, v_3 : vec2, v_4 : vec4) { + var input : PS_INPUT; + var param : PS_INPUT; + input.Position = v_2; + input.UV = v_3; + input.Color = v_4; + param = input; + v_1 = v_5(&(param)); +} + +fn v_5(input : ptr) -> vec4 { + var c : vec4; + c = ((*(input)).Color + v.offset); + c.w = 1.0f; + return c; +} + +@fragment +fn main(@builtin(position) v_6 : vec4, @location(0u) v_7 : vec2, @location(1u) v_8 : vec4) -> @location(0u) vec4 { + main_inner(v_6, v_7, v_8); + return v_1; +} diff --git a/src_test/Shaders/WebGPU/simple_constant_rectangle.vert b/src_test/Shaders/WebGPU/simple_constant_rectangle.vert new file mode 100644 index 00000000..98b572b0 --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_constant_rectangle.vert @@ -0,0 +1,63 @@ +diagnostic(off, derivative_uniformity); + +struct CB { + offset : vec4, +} + +@group(0) @binding(0) var v : CB; + +var v_1 : vec4; + +var v_2 : vec2; + +var v_3 : vec4; + +struct VS_INPUT { + Position : vec3, + UV : vec2, + Color : vec4, +} + +struct VS_OUTPUT { + Position : vec4, + UV : vec2, + Color : vec4, +} + +fn main_inner(v_4 : vec3, v_5 : vec2, v_6 : vec4) { + var input : VS_INPUT; + var flattenTemp : VS_OUTPUT; + var param : VS_INPUT; + input.Position = v_4; + input.UV = v_5; + input.Color = v_6; + param = input; + flattenTemp = v_7(&(param)); + v_1 = flattenTemp.Position; + v_2 = flattenTemp.UV; + v_3 = flattenTemp.Color; +} + +fn v_7(input : ptr) -> VS_OUTPUT { + var output : VS_OUTPUT; + let v_8 = (*(input)).Position; + output.Position = (vec4(v_8.x, v_8.y, v_8.z, 1.0f) + v.offset); + output.UV = (*(input)).UV; + output.Color = (*(input)).Color; + return output; +} + +struct tint_symbol { + @builtin(position) + m : vec4, + @location(0u) + m_1 : vec2, + @location(1u) + m_2 : vec4, +} + +@vertex +fn main(@location(0u) v_9 : vec3, @location(1u) v_10 : vec2, @location(2u) v_11 : vec4) -> tint_symbol { + main_inner(v_9, v_10, v_11); + return tint_symbol(v_1, v_2, v_3); +} diff --git a/src_test/Shaders/WebGPU/simple_mrt_texture_rectangle.frag b/src_test/Shaders/WebGPU/simple_mrt_texture_rectangle.frag new file mode 100644 index 00000000..6cf29608 --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_mrt_texture_rectangle.frag @@ -0,0 +1,59 @@ +diagnostic(off, derivative_uniformity); + +@group(1) @binding(0) var txt : texture_2d; + +@group(2) @binding(0) var smp : sampler; + +var v : vec4; + +var v_1 : vec4; + +struct PS_INPUT { + Position : vec4, + UV : vec2, + Color : vec4, +} + +struct PS_OUTPUT { + Color0 : vec4, + Color1 : vec4, +} + +fn main_inner(v_2 : vec4, v_3 : vec2, v_4 : vec4) { + var input : PS_INPUT; + var flattenTemp : PS_OUTPUT; + var param : PS_INPUT; + input.Position = v_2; + input.UV = v_3; + input.Color = v_4; + param = input; + flattenTemp = v_5(&(param)); + v = flattenTemp.Color0; + v_1 = flattenTemp.Color1; +} + +fn v_5(input : ptr) -> PS_OUTPUT { + var c : vec4; + var output : PS_OUTPUT; + c = textureSample(txt, smp, (*(input)).UV); + c.w = 255.0f; + output.Color0 = c; + c.x = (1.0f - c.x); + c.y = (1.0f - c.y); + c.z = (1.0f - c.z); + output.Color1 = c; + return output; +} + +struct tint_symbol { + @location(0u) + m : vec4, + @location(1u) + m_1 : vec4, +} + +@fragment +fn main(@builtin(position) v_6 : vec4, @location(0u) v_7 : vec2, @location(1u) v_8 : vec4) -> tint_symbol { + main_inner(v_6, v_7, v_8); + return tint_symbol(v, v_1); +} diff --git a/src_test/Shaders/WebGPU/simple_rectangle.frag b/src_test/Shaders/WebGPU/simple_rectangle.frag new file mode 100644 index 00000000..192cdc4b --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_rectangle.frag @@ -0,0 +1,27 @@ +diagnostic(off, derivative_uniformity); + +var v : vec4; + +struct PS_INPUT { + g_position : vec4, + g_color : vec4, +} + +fn main_inner(v_1 : vec4, v_2 : vec4) { + var input : PS_INPUT; + var param : PS_INPUT; + input.g_position = v_1; + input.g_color = v_2; + param = input; + v = v_3(&(param)); +} + +fn v_3(input : ptr) -> vec4 { + return (*(input)).g_color; +} + +@fragment +fn main(@builtin(position) v_4 : vec4, @location(0u) v_5 : vec4) -> @location(0u) vec4 { + main_inner(v_4, v_5); + return v; +} diff --git a/src_test/Shaders/WebGPU/simple_rectangle.vert b/src_test/Shaders/WebGPU/simple_rectangle.vert new file mode 100644 index 00000000..381b6d9a --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_rectangle.vert @@ -0,0 +1,50 @@ +diagnostic(off, derivative_uniformity); + +var v : vec4; + +var v_1 : vec4; + +struct VS_INPUT { + g_position : vec3, + g_uv : vec2, + g_color : vec4, +} + +struct VS_OUTPUT { + g_position : vec4, + g_color : vec4, +} + +fn main_inner(v_2 : vec3, v_3 : vec2, v_4 : vec4) { + var input : VS_INPUT; + var flattenTemp : VS_OUTPUT; + var param : VS_INPUT; + input.g_position = v_2; + input.g_uv = v_3; + input.g_color = v_4; + param = input; + flattenTemp = v_5(&(param)); + v = flattenTemp.g_position; + v_1 = flattenTemp.g_color; +} + +fn v_5(input : ptr) -> VS_OUTPUT { + var output : VS_OUTPUT; + let v_6 = (*(input)).g_position; + output.g_position = vec4(v_6.x, v_6.y, v_6.z, 1.0f); + output.g_color = (*(input)).g_color; + return output; +} + +struct tint_symbol { + @builtin(position) + m : vec4, + @location(0u) + m_1 : vec4, +} + +@vertex +fn main(@location(0u) v_7 : vec3, @location(1u) v_8 : vec2, @location(2u) v_9 : vec4) -> tint_symbol { + main_inner(v_7, v_8, v_9); + return tint_symbol(v, v_1); +} diff --git a/src_test/Shaders/WebGPU/simple_texture_rectangle.frag b/src_test/Shaders/WebGPU/simple_texture_rectangle.frag new file mode 100644 index 00000000..963fb6e5 --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_texture_rectangle.frag @@ -0,0 +1,36 @@ +diagnostic(off, derivative_uniformity); + +@group(1) @binding(0) var txt : texture_2d; + +@group(2) @binding(0) var smp : sampler; + +var v : vec4; + +struct PS_INPUT { + Position : vec4, + UV : vec2, + Color : vec4, +} + +fn main_inner(v_1 : vec4, v_2 : vec2, v_3 : vec4) { + var input : PS_INPUT; + var param : PS_INPUT; + input.Position = v_1; + input.UV = v_2; + input.Color = v_3; + param = input; + v = v_4(&(param)); +} + +fn v_4(input : ptr) -> vec4 { + var c : vec4; + c = textureSample(txt, smp, (*(input)).UV); + c.w = 255.0f; + return c; +} + +@fragment +fn main(@builtin(position) v_5 : vec4, @location(0u) v_6 : vec2, @location(1u) v_7 : vec4) -> @location(0u) vec4 { + main_inner(v_5, v_6, v_7); + return v; +} diff --git a/src_test/Shaders/WebGPU/simple_texture_rectangle.vert b/src_test/Shaders/WebGPU/simple_texture_rectangle.vert new file mode 100644 index 00000000..1880349c --- /dev/null +++ b/src_test/Shaders/WebGPU/simple_texture_rectangle.vert @@ -0,0 +1,57 @@ +diagnostic(off, derivative_uniformity); + +var v : vec4; + +var v_1 : vec2; + +var v_2 : vec4; + +struct VS_INPUT { + Position : vec3, + UV : vec2, + Color : vec4, +} + +struct VS_OUTPUT { + Position : vec4, + UV : vec2, + Color : vec4, +} + +fn main_inner(v_3 : vec3, v_4 : vec2, v_5 : vec4) { + var input : VS_INPUT; + var flattenTemp : VS_OUTPUT; + var param : VS_INPUT; + input.Position = v_3; + input.UV = v_4; + input.Color = v_5; + param = input; + flattenTemp = v_6(&(param)); + v = flattenTemp.Position; + v_1 = flattenTemp.UV; + v_2 = flattenTemp.Color; +} + +fn v_6(input : ptr) -> VS_OUTPUT { + var output : VS_OUTPUT; + let v_7 = (*(input)).Position; + output.Position = vec4(v_7.x, v_7.y, v_7.z, 1.0f); + output.UV = (*(input)).UV; + output.Color = (*(input)).Color; + return output; +} + +struct tint_symbol { + @builtin(position) + m : vec4, + @location(0u) + m_1 : vec2, + @location(1u) + m_2 : vec4, +} + +@vertex +fn main(@location(0u) v_8 : vec3, @location(1u) v_9 : vec2, @location(2u) v_10 : vec4) -> tint_symbol { + main_inner(v_8, v_9, v_10); + return tint_symbol(v, v_1, v_2); +} diff --git a/src_test/Shaders/WebGPU/textures.frag b/src_test/Shaders/WebGPU/textures.frag new file mode 100644 index 00000000..e1a0e10f --- /dev/null +++ b/src_test/Shaders/WebGPU/textures.frag @@ -0,0 +1,47 @@ +diagnostic(off, derivative_uniformity); + +@group(1) @binding(0) var g_texture1 : texture_2d; + +@group(2) @binding(0) var g_sampler1 : sampler; + +@group(1) @binding(1) var g_texture2 : texture_2d_array; + +@group(2) @binding(1) var g_sampler2 : sampler; + +@group(1) @binding(2) var g_texture3 : texture_3d; + +@group(2) @binding(2) var g_sampler3 : sampler; + +var v : vec4; + +struct PS_Input { + Pos : vec4, + UV : vec2, + Color : vec4, +} + +fn main_inner(v_1 : vec4, v_2 : vec2, v_3 : vec4) { + var Input : PS_Input; + Input.Pos = v_1; + Input.UV = v_2; + Input.Color = v_3; + v = v_4(Input); +} + +fn v_4(Input : PS_Input) -> vec4 { + if ((Input.UV.x < 0.30000001192092895508f)) { + return textureSample(g_texture1, g_sampler1, Input.UV); + } else if ((Input.UV.x < 0.60000002384185791016f)) { + let v_5 = Input.UV; + let v_6 = vec3(v_5.x, v_5.y, 1.0f); + return textureSample(g_texture2, g_sampler2, v_6.xy, i32(v_6.z)); + } + let v_7 = Input.UV; + return textureSample(g_texture3, g_sampler3, vec3(v_7.x, v_7.y, 0.5f)); +} + +@fragment +fn main(@builtin(position) v_8 : vec4, @location(0u) v_9 : vec2, @location(1u) v_10 : vec4) -> @location(0u) vec4 { + main_inner(v_8, v_9, v_10); + return v; +} diff --git a/src_test/Shaders/WebGPU/vertex_structured.vert b/src_test/Shaders/WebGPU/vertex_structured.vert new file mode 100644 index 00000000..45fecc70 --- /dev/null +++ b/src_test/Shaders/WebGPU/vertex_structured.vert @@ -0,0 +1,67 @@ +diagnostic(off, derivative_uniformity); + +struct CS_INPUT { + value1 : f32, + value2 : f32, +} + +struct read_2 { + m : array, +} + +@group(1) @binding(0) var read_1 : read_2; + +var v : vec4; + +var v_1 : vec4; + +struct VS_INPUT { + g_position : vec3, + g_uv : vec2, + g_color : vec4, + InstanceId : u32, +} + +struct VS_OUTPUT { + g_position : vec4, + g_color : vec4, +} + +fn main_inner(v_2 : vec3, v_3 : vec2, v_4 : vec4, v_5 : u32) { + var input : VS_INPUT; + var flattenTemp : VS_OUTPUT; + var param : VS_INPUT; + input.g_position = v_2; + input.g_uv = v_3; + input.g_color = v_4; + input.InstanceId = v_5; + param = input; + flattenTemp = v_6(&(param)); + v = flattenTemp.g_position; + v_1 = flattenTemp.g_color; +} + +fn v_6(input : ptr) -> VS_OUTPUT { + var output : VS_OUTPUT; + let v_7 = (*(input)).g_position; + output.g_position = vec4(v_7.x, v_7.y, v_7.z, 1.0f); + let v_8 = read_1.m[(*(input)).InstanceId].value1; + output.g_position.x = (output.g_position.x + v_8); + let v_9 = read_1.m[(*(input)).InstanceId].value2; + output.g_position.y = (output.g_position.y + v_9); + output.g_color = (*(input)).g_color; + return output; +} + +struct tint_symbol { + @builtin(position) + m_1 : vec4, + @location(0u) + m_2 : vec4, +} + +@vertex +fn main(@location(0u) v_10 : vec3, @location(1u) v_11 : vec2, @location(2u) v_12 : vec4, @builtin(instance_index) v_13 : u32) -> tint_symbol { + main_inner(v_10, v_11, v_12, v_13); + return tint_symbol(v, v_1); +} diff --git a/src_test/Shaders/WebGPU/vtf.vert b/src_test/Shaders/WebGPU/vtf.vert new file mode 100644 index 00000000..3ba83595 --- /dev/null +++ b/src_test/Shaders/WebGPU/vtf.vert @@ -0,0 +1,60 @@ +diagnostic(off, derivative_uniformity); + +@group(1) @binding(0) var txt : texture_2d; + +@group(2) @binding(0) var smp : sampler; + +var v : vec4; + +var v_1 : vec4; + +struct VS_INPUT { + g_position : vec3, + g_uv : vec2, + g_color : vec4, +} + +struct VS_OUTPUT { + g_position : vec4, + g_color : vec4, +} + +fn main_inner(v_2 : vec3, v_3 : vec2, v_4 : vec4) { + var input : VS_INPUT; + var flattenTemp : VS_OUTPUT; + var param : VS_INPUT; + input.g_position = v_2; + input.g_uv = v_3; + input.g_color = v_4; + param = input; + flattenTemp = v_5(&(param)); + v = flattenTemp.g_position; + v_1 = flattenTemp.g_color; +} + +fn v_5(input : ptr) -> VS_OUTPUT { + var c : vec4; + var output : VS_OUTPUT; + c = textureSampleLevel(txt, smp, (*(input)).g_uv, 0.0f); + let v_6 = (*(input)).g_position; + output.g_position = vec4(v_6.x, v_6.y, v_6.z, 1.0f); + let v_7 = c.xy; + let v_8 = (output.g_position.xy + v_7); + output.g_position.x = v_8.x; + output.g_position.y = v_8.y; + output.g_color = (*(input)).g_color; + return output; +} + +struct tint_symbol { + @builtin(position) + m : vec4, + @location(0u) + m_1 : vec4, +} + +@vertex +fn main(@location(0u) v_9 : vec3, @location(1u) v_10 : vec2, @location(2u) v_11 : vec4) -> tint_symbol { + main_inner(v_9, v_10, v_11); + return tint_symbol(v, v_1); +} diff --git a/src_test/Shaders/WebGPU_Compiled/basic.comp b/src_test/Shaders/WebGPU_Compiled/basic.comp new file mode 100644 index 00000000..3c25587f Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/basic.comp differ diff --git a/src_test/Shaders/WebGPU_Compiled/instancing.vert b/src_test/Shaders/WebGPU_Compiled/instancing.vert new file mode 100644 index 00000000..d8250d1c Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/instancing.vert differ diff --git a/src_test/Shaders/WebGPU_Compiled/readwrite.comp b/src_test/Shaders/WebGPU_Compiled/readwrite.comp new file mode 100644 index 00000000..1aefab54 Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/readwrite.comp differ diff --git a/src_test/Shaders/WebGPU_Compiled/readwrite_texture.comp b/src_test/Shaders/WebGPU_Compiled/readwrite_texture.comp new file mode 100644 index 00000000..619397e0 Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/readwrite_texture.comp differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_compute_rectangle.frag b/src_test/Shaders/WebGPU_Compiled/simple_compute_rectangle.frag new file mode 100644 index 00000000..32cf9f57 Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_compute_rectangle.frag differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_compute_rectangle.vert b/src_test/Shaders/WebGPU_Compiled/simple_compute_rectangle.vert new file mode 100644 index 00000000..e8a33fdd Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_compute_rectangle.vert differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_constant_rectangle.frag b/src_test/Shaders/WebGPU_Compiled/simple_constant_rectangle.frag new file mode 100644 index 00000000..e1ee212d Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_constant_rectangle.frag differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_constant_rectangle.vert b/src_test/Shaders/WebGPU_Compiled/simple_constant_rectangle.vert new file mode 100644 index 00000000..dff47a7d Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_constant_rectangle.vert differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_mrt_texture_rectangle.frag b/src_test/Shaders/WebGPU_Compiled/simple_mrt_texture_rectangle.frag new file mode 100644 index 00000000..c82fd5cb Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_mrt_texture_rectangle.frag differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_rectangle.frag b/src_test/Shaders/WebGPU_Compiled/simple_rectangle.frag new file mode 100644 index 00000000..33e15027 Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_rectangle.frag differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_rectangle.vert b/src_test/Shaders/WebGPU_Compiled/simple_rectangle.vert new file mode 100644 index 00000000..f064f15d Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_rectangle.vert differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_texture_rectangle.frag b/src_test/Shaders/WebGPU_Compiled/simple_texture_rectangle.frag new file mode 100644 index 00000000..2fbf32f8 Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_texture_rectangle.frag differ diff --git a/src_test/Shaders/WebGPU_Compiled/simple_texture_rectangle.vert b/src_test/Shaders/WebGPU_Compiled/simple_texture_rectangle.vert new file mode 100644 index 00000000..e42e060b Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/simple_texture_rectangle.vert differ diff --git a/src_test/Shaders/WebGPU_Compiled/textures.frag b/src_test/Shaders/WebGPU_Compiled/textures.frag new file mode 100644 index 00000000..f686aefd Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/textures.frag differ diff --git a/src_test/Shaders/WebGPU_Compiled/vertex_structured.vert b/src_test/Shaders/WebGPU_Compiled/vertex_structured.vert new file mode 100644 index 00000000..0643f089 Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/vertex_structured.vert differ diff --git a/src_test/Shaders/WebGPU_Compiled/vtf.vert b/src_test/Shaders/WebGPU_Compiled/vtf.vert new file mode 100644 index 00000000..85dd5fb9 Binary files /dev/null and b/src_test/Shaders/WebGPU_Compiled/vtf.vert differ diff --git a/src_test/TestHelper.cpp b/src_test/TestHelper.cpp index bf60cdf9..ad542b31 100644 --- a/src_test/TestHelper.cpp +++ b/src_test/TestHelper.cpp @@ -32,12 +32,31 @@ class DefaultTestFileReader : public TestFileReader return ret; } - fseek(fp, 0, SEEK_END); - auto size = ftell(fp); - fseek(fp, 0, SEEK_SET); + if (fseek(fp, 0, SEEK_END) != 0) + { + fclose(fp); + return ret; + } + + const auto size = ftell(fp); + if (size <= 0) + { + fclose(fp); + return ret; + } - ret.resize(size); - fread(ret.data(), 1, size, fp); + if (fseek(fp, 0, SEEK_SET) != 0) + { + fclose(fp); + return ret; + } + + ret.resize(static_cast(size)); + const auto readSize = fread(ret.data(), 1, ret.size(), fp); + if (readSize != ret.size()) + { + ret.resize(readSize); + } fclose(fp); return ret; @@ -71,6 +90,7 @@ ParsedArgs TestHelper::ParseArg(int argc, char* argv[]) ParsedArgs args; bool isVulkanMode = false; + bool isWebGPUMode = false; std::string filter; for (int i = 0; i < argc; i++) @@ -81,6 +101,10 @@ ParsedArgs TestHelper::ParseArg(int argc, char* argv[]) { isVulkanMode = true; } + else if (v == "--webgpu") + { + isWebGPUMode = true; + } else if (v.find("--filter=") == 0) { args.Filter = v.substr(strlen("--filter=")); @@ -99,6 +123,10 @@ ParsedArgs TestHelper::ParseArg(int argc, char* argv[]) { args.Device = LLGI::DeviceType::Vulkan; } + if (isWebGPUMode) + { + args.Device = LLGI::DeviceType::WebGPU; + } return args; } @@ -435,6 +463,10 @@ std::string TestHelper::GetDeviceName(LLGI::DeviceType device) { return std::string{"Vulkan"}; } + else if (device == LLGI::DeviceType::WebGPU) + { + return std::string{"WebGPU"}; + } return std::string{"Unknown"}; } diff --git a/src_test/browser/pre_webgpu_test.js b/src_test/browser/pre_webgpu_test.js new file mode 100644 index 00000000..34a3f06d --- /dev/null +++ b/src_test/browser/pre_webgpu_test.js @@ -0,0 +1,61 @@ +if (typeof Module === 'undefined') { + var Module = {}; +} + +Module.arguments = Module.arguments || (function() { + var args = ['--webgpu', '--filter=WebGPUBrowser.*']; + if (typeof URLSearchParams !== 'undefined' && typeof location !== 'undefined') { + var params = new URLSearchParams(location.search); + var filter = params.get('filter'); + if (filter) { + args[1] = '--filter=' + filter; + } + } + return args; +})(); + +Module.preRun = Module.preRun || []; +Module.preRun.push(function() { + var dependency = 'llgi-webgpu-device'; + addRunDependency(dependency); + + (async function() { + if (!navigator.gpu) { + throw new Error('navigator.gpu is not available. Serve over localhost/https and use a WebGPU-capable browser.'); + } + + var adapter = await navigator.gpu.requestAdapter(); + if (!adapter) { + throw new Error('Failed to request a WebGPU adapter.'); + } + + var optionalFeatures = ['float32-filterable', 'texture-formats-tier2']; + var requiredFeatures = optionalFeatures.filter(function(feature) { + return adapter.features && adapter.features.has(feature); + }); + + Module.preinitializedWebGPUDevice = await adapter.requestDevice({ + requiredFeatures: requiredFeatures + }); + Module.preinitializedWebGPUDevice.addEventListener('uncapturederror', function(event) { + Module.llgiLastWebGPUError = event.error && event.error.message ? event.error.message : String(event.error); + console.error('LLGI_WEBGPU_ERROR', Module.llgiLastWebGPUError); + }); + removeRunDependency(dependency); + })().catch(function(error) { + Module.llgiTestResult = { + status: 'failed', + message: error && error.message ? error.message : String(error) + }; + console.error('LLGI_TEST_FAIL', Module.llgiTestResult.message); + removeRunDependency(dependency); + }); +}); + +Module.onAbort = function(reason) { + Module.llgiTestResult = { + status: 'failed', + message: reason ? String(reason) : 'aborted' + }; + console.error('LLGI_TEST_FAIL', Module.llgiTestResult.message); +}; diff --git a/src_test/browser/run_webgpu_browser_test.mjs b/src_test/browser/run_webgpu_browser_test.mjs new file mode 100644 index 00000000..4963678f --- /dev/null +++ b/src_test/browser/run_webgpu_browser_test.mjs @@ -0,0 +1,111 @@ +import http from 'node:http'; +import fs from 'node:fs'; +import path from 'node:path'; + +let chromium; +try { + ({chromium} = await import('playwright')); +} catch (error) { + console.error('The "playwright" package is required. Install it with: npm install playwright && npx playwright install chromium'); + process.exit(2); +} + +const args = process.argv.slice(2); +const htmlPath = args[0]; +const filterArg = args.find((arg) => arg.startsWith('--filter=')); +const filter = filterArg ? filterArg.substring('--filter='.length) : 'WebGPUBrowser.*'; + +if (!htmlPath) { + console.error('Usage: node run_webgpu_browser_test.mjs [--filter=WebGPUBrowser.*]'); + process.exit(2); +} + +const resolvedHtmlPath = path.resolve(htmlPath); +if (!fs.existsSync(resolvedHtmlPath)) { + console.error(`Not found: ${resolvedHtmlPath}`); + process.exit(2); +} + +const root = path.dirname(resolvedHtmlPath); +const htmlFile = path.basename(resolvedHtmlPath); +const executablePath = process.env.CHROME_PATH || process.env.PLAYWRIGHT_CHROMIUM_EXECUTABLE_PATH; + +function contentType(filePath) { + if (filePath.endsWith('.html')) return 'text/html'; + if (filePath.endsWith('.js')) return 'application/javascript'; + if (filePath.endsWith('.wasm')) return 'application/wasm'; + if (filePath.endsWith('.data')) return 'application/octet-stream'; + return 'application/octet-stream'; +} + +const server = http.createServer((request, response) => { + const requestUrl = new URL(request.url, 'http://127.0.0.1'); + if (requestUrl.pathname === '/favicon.ico') { + response.writeHead(204); + response.end(); + return; + } + + const relativePath = decodeURIComponent(requestUrl.pathname === '/' ? `/${htmlFile}` : requestUrl.pathname); + const filePath = path.resolve(root, `.${relativePath}`); + + if (!filePath.startsWith(root) || !fs.existsSync(filePath) || !fs.statSync(filePath).isFile()) { + response.writeHead(404); + response.end('Not found'); + return; + } + + response.writeHead(200, { + 'Content-Type': contentType(filePath), + 'Cache-Control': 'no-store, max-age=0', + 'Pragma': 'no-cache', + 'Expires': '0', + }); + fs.createReadStream(filePath).pipe(response); +}); + +await new Promise((resolve) => server.listen(0, '127.0.0.1', resolve)); +const {port} = server.address(); +const url = `http://127.0.0.1:${port}/${htmlFile}?filter=${encodeURIComponent(filter)}`; + +let browser; +try { + browser = await chromium.launch({ + headless: true, + executablePath, + args: [ + '--enable-unsafe-webgpu', + '--ignore-gpu-blocklist', + '--enable-features=Vulkan,UseSkiaRenderer', + '--use-vulkan=swiftshader', + ], + }); + + const page = await browser.newPage(); + page.on('console', (message) => console.log(`[browser:${message.type()}] ${message.text()}`)); + page.on('pageerror', (error) => console.error(`[browser:pageerror] ${error.message}`)); + + await page.goto(url, {waitUntil: 'load'}); + const result = await page.waitForFunction( + () => globalThis.Module && globalThis.Module.llgiTestResult, + null, + {timeout: 60000} + ); + const value = await result.jsonValue(); + if (!value || value.status !== 'passed') { + console.error(value && value.message ? value.message : 'LLGI browser WebGPU test failed.'); + process.exitCode = 1; + } else { + await page.waitForTimeout(500); + const lateError = await page.evaluate(() => globalThis.Module && globalThis.Module.llgiLastWebGPUError); + if (lateError) { + console.error(lateError); + process.exitCode = 1; + } + } +} finally { + if (browser) { + await browser.close(); + } + await new Promise((resolve) => server.close(resolve)); +} diff --git a/src_test/capture.cpp b/src_test/capture.cpp index beb0043e..46145cff 100644 --- a/src_test/capture.cpp +++ b/src_test/capture.cpp @@ -175,7 +175,7 @@ void test_capture(LLGI::DeviceType deviceType, LLGI::Vec2I windowSize) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); diff --git a/src_test/main.cpp b/src_test/main.cpp index ace89923..19782a0e 100644 --- a/src_test/main.cpp +++ b/src_test/main.cpp @@ -5,6 +5,21 @@ #include #include +#ifdef __EMSCRIPTEN__ +#include + +EM_JS(void, llgi_report_test_result, (int result, const char* message), { + let text = UTF8ToString(message); + let status = result === 0 ? 'passed' : 'failed'; + if (status === 'passed' && Module.llgiLastWebGPUError) { + status = 'failed'; + text = Module.llgiLastWebGPUError; + } + Module.llgiTestResult = { status: status, message: text }; + console.log(status === 'passed' ? 'LLGI_TEST_PASS' : 'LLGI_TEST_FAIL', text); +}); +#endif + #ifdef _WIN32 #pragma comment(lib, "d3dcompiler.lib") @@ -14,7 +29,7 @@ #endif -#if defined(__linux__) || defined(__APPLE__) || defined(_WIN32) +#if defined(__linux__) || defined(__APPLE__) || defined(_WIN32) || defined(__EMSCRIPTEN__) int main(int argc, char* argv[]) { @@ -28,6 +43,9 @@ int main(int argc, char* argv[]) // make shaders folder path from __FILE__ { +#if defined(__EMSCRIPTEN__) + TestHelper::SetRoot("/Shaders/WebGPU/"); +#else auto path = std::string(__FILE__); #if defined(WIN32) auto pos = path.find_last_of("\\"); @@ -53,6 +71,11 @@ int main(int argc, char* argv[]) TestHelper::SetRoot((path + "/Shaders/SPIRV/").c_str()); #endif } + else if (args.Device == LLGI::DeviceType::WebGPU) + { + TestHelper::SetRoot((path + "/Shaders/WebGPU/").c_str()); + } +#endif } LLGI::SetLogger([](LLGI::LogType logType, const std::string& message) { std::cerr << message << std::endl; }); @@ -65,6 +88,10 @@ int main(int argc, char* argv[]) TestHelper::Dispose(); +#ifdef __EMSCRIPTEN__ + llgi_report_test_result(0, "completed"); +#endif + return 0; } #endif diff --git a/src_test/test_clear.cpp b/src_test/test_clear.cpp index eeea9039..ccd35d0b 100644 --- a/src_test/test_clear.cpp +++ b/src_test/test_clear.cpp @@ -1,6 +1,6 @@ #include "TestHelper.h" #include "test.h" -#include +#include void test_clear_update(LLGI::DeviceType deviceType) { @@ -15,9 +15,7 @@ void test_clear_update(LLGI::DeviceType deviceType) auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); while (count < 60) { @@ -32,8 +30,7 @@ void test_clear_update(LLGI::DeviceType deviceType) color.B = 0; color.A = 255; - auto commandList = commandLists[count % commandLists.size()]; - commandList->WaitUntilCompleted(); + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass( @@ -48,7 +45,7 @@ void test_clear_update(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(color, true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -61,8 +58,6 @@ void test_clear_update(LLGI::DeviceType deviceType) graphics->WaitFinish(); LLGI::SafeRelease(sfMemoryPool); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); } @@ -80,9 +75,7 @@ void test_clear(LLGI::DeviceType deviceType) auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); LLGI::Color8 color; color.R = 255; @@ -99,8 +92,7 @@ void test_clear(LLGI::DeviceType deviceType) // It need to create a command buffer between NewFrame and Present. // Because get current screen returns other values by every frame. - auto commandList = commandLists[count % commandLists.size()]; - commandList->WaitUntilCompleted(); + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass( @@ -115,7 +107,7 @@ void test_clear(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(color, true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -128,8 +120,6 @@ void test_clear(LLGI::DeviceType deviceType) graphics->WaitFinish(); LLGI::SafeRelease(sfMemoryPool); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); } diff --git a/src_test/test_compute_shader.cpp b/src_test/test_compute_shader.cpp index 95d65f92..31a8a019 100644 --- a/src_test/test_compute_shader.cpp +++ b/src_test/test_compute_shader.cpp @@ -196,6 +196,10 @@ void test_compute_shader_texture(LLGI::DeviceType deviceType) LLGI::TextureParameter texParamWrite; texParamWrite.Size = {1, 1, 1}; texParamWrite.Usage = LLGI::TextureUsageType::Storage; + if (deviceType == LLGI::DeviceType::WebGPU) + { + texParamWrite.Format = LLGI::TextureFormatType::R32G32B32A32_FLOAT; + } auto texWrite = LLGI::CreateSharedPtr(graphics->CreateTexture(texParamWrite)); if (!platform->NewFrame()) @@ -220,7 +224,16 @@ void test_compute_shader_texture(LLGI::DeviceType deviceType) std::vector result; if (texWrite->GetData(result)) { - if (!(result[0] == 128 && result[1] == 64 && result[2] == 64 && result[3] == 128)) + if (deviceType == LLGI::DeviceType::WebGPU) + { + const auto p = reinterpret_cast(result.data()); + if (!(p[0] == 0.5f && p[1] == 0.25f && p[2] == 0.25f && p[3] == 0.5f)) + { + std::cout << "Failed : Mismatch" << p[0] << "," << p[1] << "," << p[2] << "," << p[3] << std::endl; + abort(); + } + } + else if (!(result[0] == 128 && result[1] == 64 && result[2] == 64 && result[3] == 128)) { std::cout << "Failed : Mismatch" << static_cast(result[0]) << "," << static_cast(result[1]) << "," diff --git a/src_test/test_mipmap.cpp b/src_test/test_mipmap.cpp index ade5f4ce..c7d855e9 100644 --- a/src_test/test_mipmap.cpp +++ b/src_test/test_mipmap.cpp @@ -3,7 +3,6 @@ #include "test.h" #include -#include #include #include #include @@ -23,9 +22,7 @@ void test_mipmap(LLGI::DeviceType deviceType) auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); LLGI::TextureInitializationParameter texParam_mipmap; @@ -76,7 +73,7 @@ void test_mipmap(LLGI::DeviceType deviceType) LLGI::CompilerResult result_ps; if (platform->GetDeviceType() == LLGI::DeviceType::Metal || platform->GetDeviceType() == LLGI::DeviceType::DirectX12 || - platform->GetDeviceType() == LLGI::DeviceType::Vulkan) + platform->GetDeviceType() == LLGI::DeviceType::Vulkan || platform->GetDeviceType() == LLGI::DeviceType::WebGPU) { auto code_vs = TestHelper::LoadData("simple_texture_rectangle.vert"); auto code_ps = TestHelper::LoadData("simple_texture_rectangle.frag"); @@ -196,7 +193,7 @@ void test_mipmap(LLGI::DeviceType deviceType) pips[renderPassPipelineState] = LLGI::CreateSharedPtr(pip); } - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->GenerateMipMap(textureDrawnMipmap); @@ -243,7 +240,7 @@ void test_mipmap(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto textureMipmap = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(textureMipmap); @@ -261,8 +258,6 @@ void test_mipmap(LLGI::DeviceType deviceType) LLGI::SafeRelease(textureDrawnMipmap); LLGI::SafeRelease(shader_vs); LLGI::SafeRelease(shader_ps); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); diff --git a/src_test/test_renderPass.cpp b/src_test/test_renderPass.cpp index 9c7ae401..1c55e0e2 100644 --- a/src_test/test_renderPass.cpp +++ b/src_test/test_renderPass.cpp @@ -1,6 +1,6 @@ #include "TestHelper.h" #include "test.h" -#include +#include #include enum class RenderPassTestMode @@ -33,9 +33,7 @@ void test_renderPass(LLGI::DeviceType deviceType, RenderPassTestMode mode) auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); LLGI::RenderTextureInitializationParameter params; params.Size = LLGI::Vec2I(256, 256); @@ -173,7 +171,7 @@ void test_renderPass(LLGI::DeviceType deviceType, RenderPassTestMode mode) color2.B = 0; color2.A = 255; - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass(renderPass); @@ -264,7 +262,7 @@ void test_renderPass(LLGI::DeviceType deviceType, RenderPassTestMode mode) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto screenTex = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(screenTex); @@ -293,10 +291,7 @@ void test_renderPass(LLGI::DeviceType deviceType, RenderPassTestMode mode) } } - for (size_t i = 0; i < commandLists.size(); i++) - { - commandLists[i]->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); graphics->WaitFinish(); pips.clear(); @@ -308,10 +303,6 @@ void test_renderPass(LLGI::DeviceType deviceType, RenderPassTestMode mode) LLGI::SafeRelease(depthTextureDst); LLGI::SafeRelease(renderPass); LLGI::SafeRelease(texture); - for (size_t i = 0; i < commandLists.size(); i++) - { - LLGI::SafeRelease(commandLists[i]); - } LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); } @@ -332,9 +323,7 @@ void test_copyTextureToScreen(LLGI::DeviceType deviceType) auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); LLGI::RenderTextureInitializationParameter params; params.Size = LLGI::Vec2I(1280, 720); @@ -369,7 +358,7 @@ void test_copyTextureToScreen(LLGI::DeviceType deviceType) color2.B = 0; color2.A = 255; - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass(renderPass); commandList->EndRenderPass(); @@ -385,7 +374,7 @@ void test_copyTextureToScreen(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -394,17 +383,12 @@ void test_copyTextureToScreen(LLGI::DeviceType deviceType) } } - for (size_t i = 0; i < commandLists.size(); i++) - { - commandLists[i]->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); graphics->WaitFinish(); LLGI::SafeRelease(sfMemoryPool); LLGI::SafeRelease(renderTexture); LLGI::SafeRelease(renderPass); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); } @@ -424,9 +408,7 @@ void test_multiRenderPass(LLGI::DeviceType deviceType) auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics.get(), sfMemoryPool, 3); LLGI::TextureInitializationParameter texParam; texParam.Size = LLGI::Vec2I(256, 256); @@ -512,7 +494,7 @@ void test_multiRenderPass(LLGI::DeviceType deviceType) color2.B = 0; color2.A = 255; - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass(renderPass); @@ -595,7 +577,7 @@ void test_multiRenderPass(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto screenTexture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(screenTexture); Bitmap2D(data, screenTexture->GetSizeAs2D().X, screenTexture->GetSizeAs2D().Y, screenTexture->GetFormat()) @@ -603,10 +585,7 @@ void test_multiRenderPass(LLGI::DeviceType deviceType) } } - for (size_t i = 0; i < commandLists.size(); i++) - { - commandLists[i]->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); graphics->WaitFinish(); pips.clear(); @@ -616,10 +595,6 @@ void test_multiRenderPass(LLGI::DeviceType deviceType) LLGI::SafeRelease(renderTexture2); LLGI::SafeRelease(renderPass); LLGI::SafeRelease(texture); - for (size_t i = 0; i < commandLists.size(); i++) - { - LLGI::SafeRelease(commandLists[i]); - } LLGI::SafeRelease(compiler); } diff --git a/src_test/test_simple_render.cpp b/src_test/test_simple_render.cpp index 9c6904ab..44aa587b 100644 --- a/src_test/test_simple_render.cpp +++ b/src_test/test_simple_render.cpp @@ -134,7 +134,7 @@ void test_simple_rectangle(LLGI::DeviceType deviceType, SingleRectangleTestMode if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -157,11 +157,7 @@ void test_simple_rectangle(LLGI::DeviceType deviceType, SingleRectangleTestMode } } - for (int i = 0; i < 3; i++) - { - auto commandList = commandListPool->Get(); - commandList->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); pips.clear(); @@ -187,9 +183,7 @@ void test_index_offset(LLGI::DeviceType deviceType) auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); std::shared_ptr shader_vs = nullptr; std::shared_ptr shader_ps = nullptr; @@ -243,7 +237,7 @@ void test_index_offset(LLGI::DeviceType deviceType) pips[renderPassPipelineState] = LLGI::CreateSharedPtr(pip); } - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass(renderPass); commandList->SetVertexBuffer(vb.get(), sizeof(SimpleVertex), 0); @@ -260,7 +254,7 @@ void test_index_offset(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); Bitmap2D(data, texture->GetSizeAs2D().X, texture->GetSizeAs2D().Y, texture->GetFormat()) @@ -269,19 +263,13 @@ void test_index_offset(LLGI::DeviceType deviceType) } } - for (size_t i = 0; i < commandLists.size(); i++) - { - auto commandList = commandLists[i]; - commandList->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); pips.clear(); graphics->WaitFinish(); LLGI::SafeRelease(sfMemoryPool); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); } @@ -354,9 +342,7 @@ void main() auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); LLGI::Buffer* cb_vs = nullptr; LLGI::Buffer* cb_ps = nullptr; @@ -392,7 +378,7 @@ void main() LLGI::CompilerResult result_ps; if (platform->GetDeviceType() == LLGI::DeviceType::Metal || platform->GetDeviceType() == LLGI::DeviceType::DirectX12 || - platform->GetDeviceType() == LLGI::DeviceType::Vulkan) + platform->GetDeviceType() == LLGI::DeviceType::Vulkan || platform->GetDeviceType() == LLGI::DeviceType::WebGPU) { auto code_vs = TestHelper::LoadData("simple_constant_rectangle.vert"); auto code_ps = TestHelper::LoadData("simple_constant_rectangle.frag"); @@ -517,7 +503,7 @@ void main() pips[renderPassPipelineState] = LLGI::CreateSharedPtr(pip); } - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass(renderPass); @@ -544,7 +530,7 @@ void main() if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -563,11 +549,7 @@ void main() } } - for (int i = 0; i < commandLists.size(); i++) - { - auto commandList = commandLists[i]; - commandList->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); graphics->WaitFinish(); pips.clear(); @@ -577,8 +559,6 @@ void main() LLGI::SafeRelease(cb_ps); LLGI::SafeRelease(shader_vs); LLGI::SafeRelease(shader_ps); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); @@ -644,9 +624,7 @@ void main() auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); LLGI::TextureInitializationParameter texParam; @@ -703,7 +681,7 @@ void main() LLGI::CompilerResult result_ps; if (platform->GetDeviceType() == LLGI::DeviceType::Metal || platform->GetDeviceType() == LLGI::DeviceType::DirectX12 || - platform->GetDeviceType() == LLGI::DeviceType::Vulkan) + platform->GetDeviceType() == LLGI::DeviceType::Vulkan || platform->GetDeviceType() == LLGI::DeviceType::WebGPU) { auto code_vs = TestHelper::LoadData("simple_texture_rectangle.vert"); auto code_ps = TestHelper::LoadData("simple_texture_rectangle.frag"); @@ -788,7 +766,7 @@ void main() pips[renderPassPipelineState] = LLGI::CreateSharedPtr(pip); } - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->BeginRenderPass(renderPass); // commandList->SetConstantBuffer(dummy_cb.get(), LLGI::ShaderStageType::Vertex); @@ -807,7 +785,7 @@ void main() if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -832,11 +810,7 @@ void main() } } - for (size_t i = 0; i < commandLists.size(); i++) - { - auto commandList = commandLists[i]; - commandList->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); graphics->WaitFinish(); pips.clear(); @@ -845,8 +819,6 @@ void main() LLGI::SafeRelease(textureDrawn); LLGI::SafeRelease(shader_vs); LLGI::SafeRelease(shader_ps); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); @@ -954,7 +926,7 @@ void test_instancing(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -964,11 +936,7 @@ void test_instancing(LLGI::DeviceType deviceType) } } - for (int i = 0; i < 3; i++) - { - auto commandList = commandListPool->Get(); - commandList->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); graphics->WaitFinish(); pips.clear(); } @@ -1092,7 +1060,7 @@ void test_vertex_structured(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -1206,7 +1174,7 @@ void test_vtf(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto texture = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(texture); @@ -1216,11 +1184,7 @@ void test_vtf(LLGI::DeviceType deviceType) } } - for (int i = 0; i < 3; i++) - { - auto commandList = commandListPool->Get(); - commandList->WaitUntilCompleted(); - } + commandListPool->WaitUntilCompleted(); graphics->WaitFinish(); pips.clear(); } diff --git a/src_test/test_textures.cpp b/src_test/test_textures.cpp index 975ffa96..428eefae 100644 --- a/src_test/test_textures.cpp +++ b/src_test/test_textures.cpp @@ -3,7 +3,6 @@ #include "test.h" #include -#include #include #include #include @@ -23,9 +22,7 @@ void test_textures(LLGI::DeviceType deviceType) auto graphics = platform->CreateGraphics(); auto sfMemoryPool = graphics->CreateSingleFrameMemoryPool(1024 * 1024, 128); - std::array commandLists; - for (size_t i = 0; i < commandLists.size(); i++) - commandLists[i] = graphics->CreateCommandList(sfMemoryPool); + auto commandListPool = std::make_shared(graphics, sfMemoryPool, 3); // Create textures LLGI::TextureInitializationParameter texParamSrc1; @@ -118,7 +115,7 @@ void test_textures(LLGI::DeviceType deviceType) LLGI::CompilerResult result_ps; if (platform->GetDeviceType() == LLGI::DeviceType::Metal || platform->GetDeviceType() == LLGI::DeviceType::DirectX12 || - platform->GetDeviceType() == LLGI::DeviceType::Vulkan) + platform->GetDeviceType() == LLGI::DeviceType::Vulkan || platform->GetDeviceType() == LLGI::DeviceType::WebGPU) { auto code_vs = TestHelper::LoadData("simple_texture_rectangle.vert"); auto code_ps = TestHelper::LoadData("textures.frag"); @@ -196,7 +193,7 @@ void test_textures(LLGI::DeviceType deviceType) pips[renderPassPipelineState] = LLGI::CreateSharedPtr(pip); } - auto commandList = commandLists[count % commandLists.size()]; + auto commandList = commandListPool->Get(); commandList->Begin(); commandList->CopyTexture(texSrc1.get(), texDst1.get(), {0, 0, 0}, {0, 0, 0}, {1, 1, 1}, 0, 0); @@ -221,7 +218,7 @@ void test_textures(LLGI::DeviceType deviceType) if (TestHelper::GetIsCaptureRequired() && count == 30) { - commandList->WaitUntilCompleted(); + commandListPool->WaitUntilCompleted(); auto textureMipmap = platform->GetCurrentScreen(LLGI::Color8(), true)->GetRenderTexture(0); auto data = graphics->CaptureRenderTarget(textureMipmap); @@ -237,8 +234,6 @@ void test_textures(LLGI::DeviceType deviceType) LLGI::SafeRelease(sfMemoryPool); LLGI::SafeRelease(shader_vs); LLGI::SafeRelease(shader_ps); - for (size_t i = 0; i < commandLists.size(); i++) - LLGI::SafeRelease(commandLists[i]); LLGI::SafeRelease(graphics); LLGI::SafeRelease(platform); diff --git a/src_test/test_webgpu_browser.cpp b/src_test/test_webgpu_browser.cpp new file mode 100644 index 00000000..af1cac16 --- /dev/null +++ b/src_test/test_webgpu_browser.cpp @@ -0,0 +1,520 @@ +#include "TestHelper.h" +#include "test.h" + +#include +#include +#include + +namespace +{ +struct BrowserComputeInput +{ + float value1; + float value2; +}; + +struct BrowserComputeOutput +{ + float value; +}; + +struct BrowserConstant +{ + float values[4]; +}; + +void configureSimpleVertexLayout(LLGI::PipelineState* pipelineState) +{ + pipelineState->VertexLayouts[0] = LLGI::VertexLayoutFormat::R32G32B32_FLOAT; + pipelineState->VertexLayouts[1] = LLGI::VertexLayoutFormat::R32G32_FLOAT; + pipelineState->VertexLayouts[2] = LLGI::VertexLayoutFormat::R8G8B8A8_UNORM; + pipelineState->VertexLayoutNames[0] = "POSITION"; + pipelineState->VertexLayoutNames[1] = "UV"; + pipelineState->VertexLayoutNames[2] = "COLOR"; + pipelineState->VertexLayoutCount = 3; +} + +void test_webgpu_browser_offscreen_render(LLGI::DeviceType deviceType) +{ + VERIFY(deviceType == LLGI::DeviceType::WebGPU); + + LLGI::PlatformParameter pp; + pp.Device = deviceType; + pp.WaitVSync = false; + + auto platform = LLGI::CreateSharedPtr(LLGI::CreatePlatform(pp, nullptr)); + VERIFY(platform != nullptr); + + auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); + VERIFY(graphics != nullptr); + + LLGI::RenderTextureInitializationParameter renderTextureParam; + renderTextureParam.Size = LLGI::Vec2I(64, 64); + renderTextureParam.Format = LLGI::TextureFormatType::R8G8B8A8_UNORM; + auto renderTexture = LLGI::CreateSharedPtr(graphics->CreateRenderTexture(renderTextureParam)); + VERIFY(renderTexture != nullptr); + + auto renderPass = LLGI::CreateSharedPtr(graphics->CreateRenderPass(renderTexture.get(), nullptr, nullptr, nullptr)); + VERIFY(renderPass != nullptr); + renderPass->SetClearColor(LLGI::Color8(32, 64, 96, 255)); + renderPass->SetIsColorCleared(true); + + std::shared_ptr shaderVS = nullptr; + std::shared_ptr shaderPS = nullptr; + TestHelper::CreateShader(graphics.get(), deviceType, "simple_rectangle.vert", "simple_rectangle.frag", shaderVS, shaderPS); + VERIFY(shaderVS != nullptr); + VERIFY(shaderPS != nullptr); + + std::shared_ptr vertexBuffer; + std::shared_ptr indexBuffer; + TestHelper::CreateRectangle(graphics.get(), + LLGI::Vec3F(-0.5f, 0.5f, 0.5f), + LLGI::Vec3F(0.5f, -0.5f, 0.5f), + LLGI::Color8(255, 255, 255, 255), + LLGI::Color8(0, 255, 0, 255), + vertexBuffer, + indexBuffer); + VERIFY(vertexBuffer != nullptr); + VERIFY(indexBuffer != nullptr); + + auto renderPassPipelineState = LLGI::CreateSharedPtr(graphics->CreateRenderPassPipelineState(renderPass.get())); + VERIFY(renderPassPipelineState != nullptr); + + auto pipelineState = LLGI::CreateSharedPtr(graphics->CreatePiplineState()); + VERIFY(pipelineState != nullptr); + configureSimpleVertexLayout(pipelineState.get()); + pipelineState->SetShader(LLGI::ShaderStageType::Vertex, shaderVS.get()); + pipelineState->SetShader(LLGI::ShaderStageType::Pixel, shaderPS.get()); + pipelineState->SetRenderPassPipelineState(renderPassPipelineState.get()); + VERIFY(pipelineState->Compile()); + + auto sfMemoryPool = LLGI::CreateSharedPtr(graphics->CreateSingleFrameMemoryPool(1024 * 1024, 16)); + VERIFY(sfMemoryPool != nullptr); + sfMemoryPool->NewFrame(); + + auto commandList = LLGI::CreateSharedPtr(graphics->CreateCommandList(sfMemoryPool.get())); + VERIFY(commandList != nullptr); + commandList->Begin(); + commandList->BeginRenderPass(renderPass.get()); + commandList->SetVertexBuffer(vertexBuffer.get(), sizeof(SimpleVertex), 0); + commandList->SetIndexBuffer(indexBuffer.get(), 2); + commandList->SetPipelineState(pipelineState.get()); + commandList->Draw(2); + commandList->EndRenderPass(); + commandList->End(); + + graphics->Execute(commandList.get()); + graphics->WaitFinish(); +} + +void test_webgpu_browser_texture_and_constant_render(LLGI::DeviceType deviceType) +{ + VERIFY(deviceType == LLGI::DeviceType::WebGPU); + + LLGI::PlatformParameter pp; + pp.Device = deviceType; + pp.WaitVSync = false; + + auto platform = LLGI::CreateSharedPtr(LLGI::CreatePlatform(pp, nullptr)); + VERIFY(platform != nullptr); + + auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); + VERIFY(graphics != nullptr); + + LLGI::TextureInitializationParameter textureParam; + textureParam.Size = LLGI::Vec2I(256, 256); + textureParam.Format = LLGI::TextureFormatType::R8G8B8A8_UNORM; + auto texture = LLGI::CreateSharedPtr(graphics->CreateTexture(textureParam)); + VERIFY(texture != nullptr); + TestHelper::WriteDummyTexture(texture.get()); + + LLGI::RenderTextureInitializationParameter renderTextureParam; + renderTextureParam.Size = LLGI::Vec2I(128, 128); + renderTextureParam.Format = LLGI::TextureFormatType::R8G8B8A8_UNORM; + auto renderTexture = LLGI::CreateSharedPtr(graphics->CreateRenderTexture(renderTextureParam)); + VERIFY(renderTexture != nullptr); + + auto renderPass = LLGI::CreateSharedPtr(graphics->CreateRenderPass(renderTexture.get(), nullptr, nullptr, nullptr)); + VERIFY(renderPass != nullptr); + renderPass->SetClearColor(LLGI::Color8(8, 16, 24, 255)); + renderPass->SetIsColorCleared(true); + + std::shared_ptr textureVS = nullptr; + std::shared_ptr texturePS = nullptr; + TestHelper::CreateShader(graphics.get(), deviceType, "simple_texture_rectangle.vert", "simple_texture_rectangle.frag", textureVS, texturePS); + VERIFY(textureVS != nullptr); + VERIFY(texturePS != nullptr); + + std::shared_ptr constantVS = nullptr; + std::shared_ptr constantPS = nullptr; + TestHelper::CreateShader(graphics.get(), deviceType, "simple_constant_rectangle.vert", "simple_constant_rectangle.frag", constantVS, constantPS); + VERIFY(constantVS != nullptr); + VERIFY(constantPS != nullptr); + + std::shared_ptr vertexBuffer; + std::shared_ptr indexBuffer; + TestHelper::CreateRectangle(graphics.get(), + LLGI::Vec3F(-0.8f, 0.8f, 0.5f), + LLGI::Vec3F(0.8f, -0.8f, 0.5f), + LLGI::Color8(255, 255, 255, 255), + LLGI::Color8(0, 255, 0, 255), + vertexBuffer, + indexBuffer); + VERIFY(vertexBuffer != nullptr); + VERIFY(indexBuffer != nullptr); + + auto renderPassPipelineState = LLGI::CreateSharedPtr(graphics->CreateRenderPassPipelineState(renderPass.get())); + VERIFY(renderPassPipelineState != nullptr); + + auto texturePipeline = LLGI::CreateSharedPtr(graphics->CreatePiplineState()); + VERIFY(texturePipeline != nullptr); + configureSimpleVertexLayout(texturePipeline.get()); + texturePipeline->SetShader(LLGI::ShaderStageType::Vertex, textureVS.get()); + texturePipeline->SetShader(LLGI::ShaderStageType::Pixel, texturePS.get()); + texturePipeline->SetRenderPassPipelineState(renderPassPipelineState.get()); + VERIFY(texturePipeline->Compile()); + + auto constantPipeline = LLGI::CreateSharedPtr(graphics->CreatePiplineState()); + VERIFY(constantPipeline != nullptr); + configureSimpleVertexLayout(constantPipeline.get()); + constantPipeline->SetShader(LLGI::ShaderStageType::Vertex, constantVS.get()); + constantPipeline->SetShader(LLGI::ShaderStageType::Pixel, constantPS.get()); + constantPipeline->SetRenderPassPipelineState(renderPassPipelineState.get()); + VERIFY(constantPipeline->Compile()); + + auto vertexConstantBuffer = LLGI::CreateSharedPtr( + graphics->CreateBuffer(LLGI::BufferUsageType::Constant | LLGI::BufferUsageType::MapWrite, sizeof(BrowserConstant))); + auto pixelConstantBuffer = LLGI::CreateSharedPtr( + graphics->CreateBuffer(LLGI::BufferUsageType::Constant | LLGI::BufferUsageType::MapWrite, sizeof(BrowserConstant))); + VERIFY(vertexConstantBuffer != nullptr); + VERIFY(pixelConstantBuffer != nullptr); + + { + auto data = static_cast(vertexConstantBuffer->Lock()); + data->values[0] = 0.15f; + data->values[1] = -0.10f; + data->values[2] = 0.0f; + data->values[3] = 0.0f; + vertexConstantBuffer->Unlock(); + } + { + auto data = static_cast(pixelConstantBuffer->Lock()); + data->values[0] = 0.05f; + data->values[1] = 0.10f; + data->values[2] = 0.15f; + data->values[3] = 0.0f; + pixelConstantBuffer->Unlock(); + } + + auto sfMemoryPool = LLGI::CreateSharedPtr(graphics->CreateSingleFrameMemoryPool(1024 * 1024, 16)); + VERIFY(sfMemoryPool != nullptr); + sfMemoryPool->NewFrame(); + + auto commandList = LLGI::CreateSharedPtr(graphics->CreateCommandList(sfMemoryPool.get())); + VERIFY(commandList != nullptr); + commandList->Begin(); + commandList->BeginRenderPass(renderPass.get()); + commandList->SetVertexBuffer(vertexBuffer.get(), sizeof(SimpleVertex), 0); + commandList->SetIndexBuffer(indexBuffer.get(), 2); + commandList->SetPipelineState(texturePipeline.get()); + commandList->SetTexture(texture.get(), LLGI::TextureWrapMode::Clamp, LLGI::TextureMinMagFilter::Linear, 0); + commandList->Draw(2); + commandList->ResetTextures(); + commandList->SetPipelineState(constantPipeline.get()); + commandList->SetConstantBuffer(vertexConstantBuffer.get(), 0); + commandList->SetConstantBuffer(pixelConstantBuffer.get(), 1); + commandList->Draw(2); + commandList->EndRenderPass(); + commandList->End(); + + graphics->Execute(commandList.get()); + graphics->WaitFinish(); + + const auto data = graphics->CaptureRenderTarget(renderTexture.get()); + VERIFY(data.size() == 128 * 128 * 4); + + const auto verifyPixel = [&data](int x, int y, uint8_t r, uint8_t g, uint8_t b, uint8_t a) { + const auto* pixel = data.data() + (x + y * 128) * 4; + const bool matched = std::abs(static_cast(pixel[0]) - static_cast(r)) <= 1 && + std::abs(static_cast(pixel[1]) - static_cast(g)) <= 1 && + std::abs(static_cast(pixel[2]) - static_cast(b)) <= 1 && + std::abs(static_cast(pixel[3]) - static_cast(a)) <= 1; + if (!matched) + { + std::cout << "Pixel mismatch at " << x << "," << y << " actual=" << static_cast(pixel[0]) << "," + << static_cast(pixel[1]) << "," << static_cast(pixel[2]) << "," << static_cast(pixel[3]) + << " expected=" << static_cast(r) << "," << static_cast(g) << "," << static_cast(b) << "," + << static_cast(a) << std::endl; + } + VERIFY(matched); + }; + + verifyPixel(64, 64, 155, 255, 188, 255); + verifyPixel(2, 2, 8, 16, 24, 255); +} + +void test_webgpu_browser_compute_compile(LLGI::DeviceType deviceType) +{ + VERIFY(deviceType == LLGI::DeviceType::WebGPU); + + LLGI::PlatformParameter pp; + pp.Device = deviceType; + pp.WaitVSync = false; + + auto platform = LLGI::CreateSharedPtr(LLGI::CreatePlatform(pp, nullptr)); + VERIFY(platform != nullptr); + + auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); + VERIFY(graphics != nullptr); + + std::shared_ptr shaderCS = nullptr; + TestHelper::CreateComputeShader(graphics.get(), deviceType, "basic.comp", shaderCS); + VERIFY(shaderCS != nullptr); + + auto pipelineState = LLGI::CreateSharedPtr(graphics->CreatePiplineState()); + VERIFY(pipelineState != nullptr); + pipelineState->SetShader(LLGI::ShaderStageType::Compute, shaderCS.get()); + VERIFY(pipelineState->Compile()); +} + +void test_webgpu_browser_compute_dispatch(LLGI::DeviceType deviceType) +{ + VERIFY(deviceType == LLGI::DeviceType::WebGPU); + + LLGI::PlatformParameter pp; + pp.Device = deviceType; + pp.WaitVSync = false; + + auto platform = LLGI::CreateSharedPtr(LLGI::CreatePlatform(pp, nullptr)); + VERIFY(platform != nullptr); + + auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); + VERIFY(graphics != nullptr); + + std::shared_ptr shaderCS = nullptr; + TestHelper::CreateComputeShader(graphics.get(), deviceType, "basic.comp", shaderCS); + VERIFY(shaderCS != nullptr); + + auto pipelineState = LLGI::CreateSharedPtr(graphics->CreatePiplineState()); + VERIFY(pipelineState != nullptr); + pipelineState->SetShader(LLGI::ShaderStageType::Compute, shaderCS.get()); + VERIFY(pipelineState->Compile()); + + const int dataSize = 32; + auto uploadBuffer = LLGI::CreateSharedPtr( + graphics->CreateBuffer(LLGI::BufferUsageType::MapWrite | LLGI::BufferUsageType::CopySrc, sizeof(BrowserComputeInput) * dataSize)); + auto inputBuffer = LLGI::CreateSharedPtr( + graphics->CreateBuffer(LLGI::BufferUsageType::ComputeWrite | LLGI::BufferUsageType::CopyDst, sizeof(BrowserComputeInput) * dataSize)); + auto outputBuffer = LLGI::CreateSharedPtr( + graphics->CreateBuffer(LLGI::BufferUsageType::ComputeWrite | LLGI::BufferUsageType::CopySrc, sizeof(BrowserComputeOutput) * dataSize)); + auto readbackTarget = LLGI::CreateSharedPtr( + graphics->CreateBuffer(LLGI::BufferUsageType::MapRead | LLGI::BufferUsageType::CopyDst, sizeof(BrowserComputeOutput) * dataSize)); + auto constantBuffer = LLGI::CreateSharedPtr( + graphics->CreateBuffer(LLGI::BufferUsageType::Constant | LLGI::BufferUsageType::MapWrite, sizeof(float))); + VERIFY(uploadBuffer != nullptr); + VERIFY(inputBuffer != nullptr); + VERIFY(outputBuffer != nullptr); + VERIFY(readbackTarget != nullptr); + VERIFY(constantBuffer != nullptr); + + { + auto data = static_cast(uploadBuffer->Lock()); + for (int i = 0; i < dataSize; i++) + { + data[i].value1 = static_cast(i + 1); + data[i].value2 = static_cast(i + 3); + } + uploadBuffer->Unlock(); + } + { + auto data = static_cast(constantBuffer->Lock()); + *data = 7.0f; + constantBuffer->Unlock(); + } + + auto sfMemoryPool = LLGI::CreateSharedPtr(graphics->CreateSingleFrameMemoryPool(1024 * 1024, 16)); + VERIFY(sfMemoryPool != nullptr); + sfMemoryPool->NewFrame(); + + auto commandList = LLGI::CreateSharedPtr(graphics->CreateCommandList(sfMemoryPool.get())); + VERIFY(commandList != nullptr); + commandList->Begin(); + commandList->CopyBuffer(uploadBuffer.get(), inputBuffer.get()); + commandList->BeginComputePass(); + commandList->SetPipelineState(pipelineState.get()); + commandList->SetComputeBuffer(inputBuffer.get(), sizeof(BrowserComputeInput), 0, false); + commandList->SetComputeBuffer(outputBuffer.get(), sizeof(BrowserComputeOutput), 1, false); + commandList->SetConstantBuffer(constantBuffer.get(), 0); + commandList->Dispatch(dataSize, 1, 1, 1, 1, 1); + commandList->EndComputePass(); + commandList->CopyBuffer(outputBuffer.get(), readbackTarget.get()); + commandList->End(); + + graphics->Execute(commandList.get()); + graphics->WaitFinish(); + + auto result = static_cast(readbackTarget->Lock()); + VERIFY(result != nullptr); + for (int i = 0; i < dataSize; i++) + { + const float expected = static_cast(i + 1) * static_cast(i + 3) + 7.0f; + VERIFY(std::fabs(result[i].value - expected) < 0.001f); + } + readbackTarget->Unlock(); +} + +void test_webgpu_browser_render_readback(LLGI::DeviceType deviceType) +{ + VERIFY(deviceType == LLGI::DeviceType::WebGPU); + + LLGI::PlatformParameter pp; + pp.Device = deviceType; + pp.WaitVSync = false; + + auto platform = LLGI::CreateSharedPtr(LLGI::CreatePlatform(pp, nullptr)); + VERIFY(platform != nullptr); + + auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); + VERIFY(graphics != nullptr); + + LLGI::RenderTextureInitializationParameter renderTextureParam; + renderTextureParam.Size = LLGI::Vec2I(16, 16); + renderTextureParam.Format = LLGI::TextureFormatType::R8G8B8A8_UNORM; + auto renderTexture = LLGI::CreateSharedPtr(graphics->CreateRenderTexture(renderTextureParam)); + VERIFY(renderTexture != nullptr); + + auto renderPass = LLGI::CreateSharedPtr(graphics->CreateRenderPass(renderTexture.get(), nullptr, nullptr, nullptr)); + VERIFY(renderPass != nullptr); + renderPass->SetClearColor(LLGI::Color8(11, 22, 33, 255)); + renderPass->SetIsColorCleared(true); + + auto sfMemoryPool = LLGI::CreateSharedPtr(graphics->CreateSingleFrameMemoryPool(1024 * 1024, 16)); + VERIFY(sfMemoryPool != nullptr); + sfMemoryPool->NewFrame(); + + auto commandList = LLGI::CreateSharedPtr(graphics->CreateCommandList(sfMemoryPool.get())); + VERIFY(commandList != nullptr); + commandList->Begin(); + commandList->BeginRenderPass(renderPass.get()); + commandList->EndRenderPass(); + commandList->End(); + + graphics->Execute(commandList.get()); + graphics->WaitFinish(); + + const auto data = graphics->CaptureRenderTarget(renderTexture.get()); + VERIFY(data.size() == 16 * 16 * 4); + for (int i = 0; i < 16 * 16; i++) + { + const auto* pixel = data.data() + i * 4; + VERIFY(pixel[0] == 11); + VERIFY(pixel[1] == 22); + VERIFY(pixel[2] == 33); + VERIFY(pixel[3] == 255); + } +} + +void test_webgpu_browser_screen_presentation(LLGI::DeviceType deviceType) +{ + VERIFY(deviceType == LLGI::DeviceType::WebGPU); + + LLGI::PlatformParameter pp; + pp.Device = deviceType; + pp.WaitVSync = false; + + auto window = std::unique_ptr(LLGI::CreateWindow("WebGPU Browser", LLGI::Vec2I(640, 360))); + VERIFY(window != nullptr); + + auto platform = LLGI::CreateSharedPtr(LLGI::CreatePlatform(pp, window.get())); + VERIFY(platform != nullptr); + + auto graphics = LLGI::CreateSharedPtr(platform->CreateGraphics()); + VERIFY(graphics != nullptr); + + auto sfMemoryPool = LLGI::CreateSharedPtr(graphics->CreateSingleFrameMemoryPool(1024 * 1024, 16)); + VERIFY(sfMemoryPool != nullptr); + + std::shared_ptr shaderVS = nullptr; + std::shared_ptr shaderPS = nullptr; + TestHelper::CreateShader(graphics.get(), deviceType, "simple_rectangle.vert", "simple_rectangle.frag", shaderVS, shaderPS); + VERIFY(shaderVS != nullptr); + VERIFY(shaderPS != nullptr); + + std::shared_ptr vertexBuffer; + std::shared_ptr indexBuffer; + TestHelper::CreateRectangle(graphics.get(), + LLGI::Vec3F(-0.6f, 0.6f, 0.5f), + LLGI::Vec3F(0.6f, -0.6f, 0.5f), + LLGI::Color8(255, 255, 255, 255), + LLGI::Color8(0, 255, 96, 255), + vertexBuffer, + indexBuffer); + VERIFY(vertexBuffer != nullptr); + VERIFY(indexBuffer != nullptr); + + std::shared_ptr renderPassPipelineState; + std::shared_ptr pipelineState; + + for (int frame = 0; frame < 3; frame++) + { + VERIFY(platform->NewFrame()); + sfMemoryPool->NewFrame(); + + const LLGI::Color8 color(28, 96, 180, 255); + auto renderPass = platform->GetCurrentScreen(color, true, false); + VERIFY(renderPass != nullptr); + if (pipelineState == nullptr) + { + renderPassPipelineState = LLGI::CreateSharedPtr(graphics->CreateRenderPassPipelineState(renderPass)); + VERIFY(renderPassPipelineState != nullptr); + + pipelineState = LLGI::CreateSharedPtr(graphics->CreatePiplineState()); + VERIFY(pipelineState != nullptr); + configureSimpleVertexLayout(pipelineState.get()); + pipelineState->SetShader(LLGI::ShaderStageType::Vertex, shaderVS.get()); + pipelineState->SetShader(LLGI::ShaderStageType::Pixel, shaderPS.get()); + pipelineState->SetRenderPassPipelineState(renderPassPipelineState.get()); + VERIFY(pipelineState->Compile()); + } + + auto commandList = LLGI::CreateSharedPtr(graphics->CreateCommandList(sfMemoryPool.get())); + VERIFY(commandList != nullptr); + commandList->Begin(); + commandList->BeginRenderPass(renderPass); + commandList->SetVertexBuffer(vertexBuffer.get(), sizeof(SimpleVertex), 0); + commandList->SetIndexBuffer(indexBuffer.get(), 2); + commandList->SetPipelineState(pipelineState.get()); + commandList->Draw(2); + commandList->EndRenderPass(); + commandList->End(); + + graphics->Execute(commandList.get()); + platform->Present(); + graphics->WaitFinish(); + } +} +} // namespace + +TestRegister WebGPUBrowser_OffscreenRender( + "WebGPUBrowser.OffscreenRender", + [](LLGI::DeviceType device) -> void { test_webgpu_browser_offscreen_render(device); }); + +TestRegister WebGPUBrowser_ComputeCompile( + "WebGPUBrowser.ComputeCompile", + [](LLGI::DeviceType device) -> void { test_webgpu_browser_compute_compile(device); }); + +TestRegister WebGPUBrowser_TextureAndConstantRender( + "WebGPUBrowser.TextureAndConstantRender", + [](LLGI::DeviceType device) -> void { test_webgpu_browser_texture_and_constant_render(device); }); + +TestRegister WebGPUBrowser_ComputeDispatch( + "WebGPUBrowser.ComputeDispatch", + [](LLGI::DeviceType device) -> void { test_webgpu_browser_compute_dispatch(device); }); + +TestRegister WebGPUBrowser_RenderReadback( + "WebGPUBrowser.RenderReadback", + [](LLGI::DeviceType device) -> void { test_webgpu_browser_render_readback(device); }); + +TestRegister WebGPUBrowser_ScreenPresentation( + "WebGPUBrowser.ScreenPresentation", + [](LLGI::DeviceType device) -> void { test_webgpu_browser_screen_presentation(device); }); diff --git a/tools/CMakeLists.txt b/tools/CMakeLists.txt index 5ba7ebb1..edfc0525 100644 --- a/tools/CMakeLists.txt +++ b/tools/CMakeLists.txt @@ -1,3 +1,7 @@ +if(BUILD_WEBGPU) + add_definitions(-DENABLE_WEBGPU) +endif() + add_subdirectory(ShaderTranspilerCore) add_subdirectory(ShaderTranspiler) install(TARGETS ShaderTranspiler DESTINATION ${CMAKE_INSTALL_BINDIR}) diff --git a/tools/ShaderTranspiler/CMakeLists.txt b/tools/ShaderTranspiler/CMakeLists.txt index 8d5627d8..e6f0f91e 100644 --- a/tools/ShaderTranspiler/CMakeLists.txt +++ b/tools/ShaderTranspiler/CMakeLists.txt @@ -3,7 +3,11 @@ project(ShaderTranspiler) add_executable(ShaderTranspiler main.cpp) -target_compile_features(ShaderTranspiler PUBLIC cxx_std_17) +if(BUILD_WEBGPU) + target_compile_features(ShaderTranspiler PUBLIC cxx_std_20) +else() + target_compile_features(ShaderTranspiler PUBLIC cxx_std_17) +endif() target_include_directories(ShaderTranspiler PUBLIC ../ShaderTranspilerCore) diff --git a/tools/ShaderTranspiler/main.cpp b/tools/ShaderTranspiler/main.cpp index 98be23e5..492e7d1d 100644 --- a/tools/ShaderTranspiler/main.cpp +++ b/tools/ShaderTranspiler/main.cpp @@ -1,5 +1,6 @@ #include +#include #include #include #include @@ -11,13 +12,13 @@ enum class OutputType VULKAN_GLSL, MSL, HLSL, + WGSL, SPV, Max, }; int main(int argc, char* argv[]) { - std::vector args; for (int i = 1; i < argc; i++) @@ -30,6 +31,7 @@ int main(int argc, char* argv[]) std::string code; std::string inputPath; std::string outputPath; + std::string compiledOutputPath; bool isES = false; bool isDX12 = false; bool plain = false; @@ -74,6 +76,11 @@ int main(int argc, char* argv[]) outputType = OutputType::VULKAN_GLSL; i += 1; } + else if (args[i] == "-W") + { + outputType = OutputType::WGSL; + i += 1; + } else if (args[i] == "-S") { outputType = OutputType::SPV; @@ -113,14 +120,14 @@ int main(int argc, char* argv[]) { if (i == args.size() - 1) { - std::cout << "Invald input" << std::endl; + std::cout << "Invald input : arg is none" << std::endl; return 0; } std::ifstream ifs(args[i + 1]); if (ifs.fail()) { - std::cout << "Invald input" << std::endl; + std::cout << "Invald input : unknown file " << args[i + 1] << std::endl; return 0; } code = std::string((std::istreambuf_iterator(ifs)), std::istreambuf_iterator()); @@ -131,7 +138,7 @@ int main(int argc, char* argv[]) { if (i == args.size() - 1) { - std::cout << "Invald output" << std::endl; + std::cout << "Invald output : arg is none" << std::endl; return 0; } @@ -139,6 +146,18 @@ int main(int argc, char* argv[]) i += 2; } + else if (args[i] == "--compiled-output") + { + if (i == args.size() - 1) + { + std::cout << "Invald compiled output : arg is none" << std::endl; + return 0; + } + + compiledOutputPath = args[i + 1]; + + i += 2; + } else { i++; @@ -180,7 +199,7 @@ int main(int argc, char* argv[]) auto generator = std::make_shared(loadFunc); - auto spirv = generator->Generate(inputPath.c_str(), code.c_str(), includeDir, macros, shaderStage, outputType == OutputType::VULKAN_GLSL); + auto spirv = generator->Generate(inputPath.c_str(), code.c_str(), includeDir, macros, shaderStage, outputType == OutputType::VULKAN_GLSL, outputType == OutputType::WGSL); if (spirv->GetData().size() == 0) { @@ -206,6 +225,10 @@ int main(int argc, char* argv[]) { transpiler = std::make_shared(shaderModel != 0 ? shaderModel : 40, isDX12); } + else if (outputType == OutputType::WGSL) + { + transpiler = std::make_shared(); + } std::cout << inputPath << " -> " << outputPath << " ShaderModel=" << shaderModel << std::endl; @@ -236,7 +259,7 @@ int main(int argc, char* argv[]) } catch (const std::runtime_error& e) { - std::cout << e.what() << std::endl; + std::cout << "Error : " << e.what() << std::endl; return 0; } @@ -247,7 +270,31 @@ int main(int argc, char* argv[]) return 0; } + if (transpiler->GetCode() == "") + { + std::cout << "No code is generated." << std::endl; + return 1; + } + outputfile << transpiler->GetCode(); + if (outputType == OutputType::WGSL && compiledOutputPath != "") + { + static const char header[] = {'w', 'g', 's', 'l', 'c', 'o', 'd', 'e'}; + const auto transpiledCode = transpiler->GetCode(); + + std::ofstream compiledOutput(compiledOutputPath, std::ios::binary); + if (compiledOutput.bad()) + { + std::cout << "Invald compiled output" << std::endl; + return 0; + } + + compiledOutput.write(header, sizeof(header)); + compiledOutput.write(transpiledCode.data(), static_cast(transpiledCode.size())); + const char terminator = 0; + compiledOutput.write(&terminator, 1); + } + return 0; } diff --git a/tools/ShaderTranspilerCore/CMakeLists.txt b/tools/ShaderTranspilerCore/CMakeLists.txt index 04180b9d..a09a6189 100644 --- a/tools/ShaderTranspilerCore/CMakeLists.txt +++ b/tools/ShaderTranspilerCore/CMakeLists.txt @@ -4,7 +4,11 @@ project(ShaderTranspilerCore) add_library( ShaderTranspilerCore STATIC ShaderTranspilerCore.cpp ShaderTranspilerCore.h) -target_compile_features(ShaderTranspilerCore PUBLIC cxx_std_17) +if(BUILD_WEBGPU) + target_compile_features(ShaderTranspilerCore PUBLIC cxx_std_20) +else() + target_compile_features(ShaderTranspilerCore PUBLIC cxx_std_17) +endif() target_include_directories(ShaderTranspilerCore PUBLIC ${LLGI_THIRDPARTY_INCLUDES}) @@ -29,7 +33,7 @@ if(USE_THIRDPARTY_DIRECTORY) endif() if(MSVC) - target_compile_options(ShaderTranspilerCore PRIVATE /W4 /WX /wd4100) + target_compile_options(ShaderTranspilerCore PRIVATE /W4 /WX /wd4100 /wd4324) else() target_compile_options(ShaderTranspilerCore PRIVATE -Wall -Werror) endif() @@ -41,3 +45,8 @@ endif() if(SPIRVCROSS_WITHOUT_INSTALL) target_compile_definitions(ShaderTranspilerCore PRIVATE ENABLE_SPIRVCROSS_WITHOUT_INSTALL) endif() + +if(BUILD_WEBGPU) + target_compile_definitions(ShaderTranspilerCore PRIVATE ENABLE_WEBGPU) + target_link_libraries(ShaderTranspilerCore PRIVATE tint_api) +endif() diff --git a/tools/ShaderTranspilerCore/ShaderTranspilerCore.cpp b/tools/ShaderTranspilerCore/ShaderTranspilerCore.cpp index db83d2dc..f482433e 100644 --- a/tools/ShaderTranspilerCore/ShaderTranspilerCore.cpp +++ b/tools/ShaderTranspilerCore/ShaderTranspilerCore.cpp @@ -25,6 +25,13 @@ #include #endif +#if (ENABLE_WEBGPU) +#include +#endif + +#include +#include + namespace LLGI { @@ -43,6 +50,50 @@ std::string Replace(std::string target, std::string from_, std::string to_) return target; } +#if (ENABLE_WEBGPU) +std::string NormalizeWGSLForLLGI(std::string code, ShaderStageType shaderStageType) +{ + for (uint32_t i = 0; i < TextureSlotMax; i++) + { + code = Replace(code, "@group(0u) @binding(" + std::to_string(300 + i) + "u)", "@group(0) @binding(" + std::to_string(i) + ")"); + code = Replace(code, "@group(0) @binding(" + std::to_string(300 + i) + ")", "@group(0) @binding(" + std::to_string(i) + ")"); + + const auto storageGroup = shaderStageType == ShaderStageType::Compute ? 2 : 1; + code = Replace(code, + "@group(0u) @binding(" + std::to_string(400 + i) + "u)", + "@group(" + std::to_string(storageGroup) + ") @binding(" + std::to_string(i) + ")"); + code = Replace(code, + "@group(0) @binding(" + std::to_string(400 + i) + ")", + "@group(" + std::to_string(storageGroup) + ") @binding(" + std::to_string(i) + ")"); + + code = Replace(code, "@group(0u) @binding(" + std::to_string(100 + i) + "u)", "@group(1) @binding(" + std::to_string(i) + ")"); + code = Replace(code, "@group(0) @binding(" + std::to_string(100 + i) + ")", "@group(1) @binding(" + std::to_string(i) + ")"); + + code = Replace(code, "@group(0u) @binding(" + std::to_string(200 + i) + "u)", "@group(1) @binding(" + std::to_string(i) + ")"); + code = Replace(code, "@group(0) @binding(" + std::to_string(200 + i) + ")", "@group(1) @binding(" + std::to_string(i) + ")"); + } + + std::stringstream input(code); + std::stringstream output; + std::string line; + while (std::getline(input, line)) + { + if (line.find(": sampler") != std::string::npos) + { + for (uint32_t i = 0; i < TextureSlotMax; i++) + { + line = Replace(line, "@group(0u) @binding(" + std::to_string(i) + "u)", "@group(2) @binding(" + std::to_string(i) + ")"); + line = Replace(line, "@group(0) @binding(" + std::to_string(i) + ")", "@group(2) @binding(" + std::to_string(i) + ")"); + } + } + output << line << "\n"; + } + code = output.str(); + + return code; +} +#endif + // https://stackoverflow.com/questions/8518743/get-directory-from-file-path-c/14631366 std::string dirnameOf(const std::string& fname) { @@ -391,6 +442,41 @@ bool SPIRVToGLSLTranspiler::Transpile(const std::shared_ptr& spirv, LLGI: return true; } +SPIRVToWGSLTranspiler::SPIRVToWGSLTranspiler() +{ +#if (ENABLE_WEBGPU) + tint::Initialize(); +#endif +} + +SPIRVToWGSLTranspiler::~SPIRVToWGSLTranspiler() +{ +#if (ENABLE_WEBGPU) + tint::Shutdown(); +#endif +} + +bool SPIRVToWGSLTranspiler::Transpile(const std::shared_ptr& spirv, LLGI::ShaderStageType shaderStageType) +{ +#if (ENABLE_WEBGPU) + tint::wgsl::writer::Options gen_options; + gen_options.allow_non_uniform_derivatives = true; + gen_options.allowed_features.features.insert(tint::wgsl::LanguageFeature::kReadonlyAndReadwriteStorageTextures); + auto result = tint::SpirvToWgsl(spirv->GetData(), gen_options); + if (result != tint::Success) + { + errorCode_ = result.Failure().reason; + return false; + } + + code_ = NormalizeWGSLForLLGI(result.Get(), shaderStageType); + return true; +#else + errorCode_ = "WGSL output requires ShaderTranspilerCore to be built with BUILD_WEBGPU=ON."; + return false; +#endif +} + class ReflectionCompiler : public spirv_cross::Compiler { public: @@ -478,7 +564,8 @@ std::shared_ptr SPIRVGenerator::Generate(const char* path, std::vector includeDirs, std::vector macros, ShaderStageType shaderStageType, - bool isYInverted) + bool isYInverted, + bool addBindingOffset) { std::string codeStr(code); glslang::TProgram program; @@ -506,11 +593,19 @@ std::shared_ptr SPIRVGenerator::Generate(const char* path, } shader.setPreamble(macro.c_str()); - // shader->setAutoMapBindings(true); - // shader->setAutoMapLocations(true); + + if (addBindingOffset) + { + shader.setShiftBinding(glslang::TResourceType::EResSampler, 0); + shader.setShiftBinding(glslang::TResourceType::EResTexture, 100); + shader.setShiftBinding(glslang::TResourceType::EResImage, 200); + shader.setShiftBinding(glslang::TResourceType::EResUbo, 300); + shader.setShiftBinding(glslang::TResourceType::EResSsbo, 400); + shader.setShiftBinding(glslang::TResourceType::EResUav, 500); + } shader.setStrings(shaderStrings, 1); - const auto messages = static_cast(EShMsgSpvRules | EShMsgVulkanRules | EShMsgReadHlsl | EShOptFull); + const auto messages = static_cast(EShMsgSpvRules | EShMsgVulkanRules | EShMsgReadHlsl | EShOptFull | EShMsgHlslOffsets); DirStackFileIncluder includer(onLoad_); includer.pushExternalLocalDirectory(dirnameOf(path)); @@ -533,6 +628,14 @@ std::shared_ptr SPIRVGenerator::Generate(const char* path, return std::make_shared(program.getInfoLog()); } + if (addBindingOffset) + { + if (!program.mapIO()) + { + return std::make_shared(program.getInfoLog()); + } + } + std::vector spirv; glslang::SpvOptions spvOptions; spvOptions.optimizeSize = true; diff --git a/tools/ShaderTranspilerCore/ShaderTranspilerCore.h b/tools/ShaderTranspilerCore/ShaderTranspilerCore.h index 77863f35..1fd77ae0 100644 --- a/tools/ShaderTranspilerCore/ShaderTranspilerCore.h +++ b/tools/ShaderTranspilerCore/ShaderTranspilerCore.h @@ -95,6 +95,14 @@ class SPIRVToGLSLTranspiler : public SPIRVTranspiler bool Transpile(const std::shared_ptr& spirv, LLGI::ShaderStageType shaderStageType) override; }; +class SPIRVToWGSLTranspiler : public SPIRVTranspiler +{ +public: + SPIRVToWGSLTranspiler(); + ~SPIRVToWGSLTranspiler() override; + bool Transpile(const std::shared_ptr& spirv, LLGI::ShaderStageType shaderStageType) override; +}; + class SPIRVReflection : public SPIRVTranspiler { public: @@ -133,7 +141,8 @@ class SPIRVGenerator std::vector includeDirs, std::vector macros, ShaderStageType shaderStageType, - bool isYInverted); + bool isYInverted, + bool addBindingOffset); }; } // namespace LLGI