Compare commits

..
Author SHA1 Message Date
Powei Feng f1a4c174c7 vk: update vulkan memory allocator lib 2025-06-12 15:23:38 -07:00
136 changed files with 23009 additions and 23174 deletions
+1 -1
View File
@@ -138,7 +138,7 @@ jobs:
- uses: actions/upload-artifact@v4
with:
name: presubmit-renderdiff-result
path: ./out/renderdiff
path: ./out/renderdiff_tests
validate-wgsl-webgpu:
name: validate-wgsl-webgpu
+2 -2
View File
@@ -31,7 +31,7 @@ repositories {
}
dependencies {
implementation 'com.google.android.filament:filament-android:1.62.0'
implementation 'com.google.android.filament:filament-android:1.61.1'
}
```
@@ -51,7 +51,7 @@ Here are all the libraries available in the group `com.google.android.filament`:
iOS projects can use CocoaPods to install the latest release:
```shell
pod 'Filament', '~> 1.62.0'
pod 'Filament', '~> 1.61.1'
```
## Documentation
+2 -4
View File
@@ -9,11 +9,9 @@ Instead, if you are authoring a PR for the main branch, add your release note to
## v1.62.0
- Add new `unfilterable` field to Filament Material's `sampler` [⚠️ **New Material Version**]
## v1.61.2
- samples: samples now have a CLI to select backend api
- materials: sampler now export their type in the material binary [⚠️ **New Material Version**]
- materials: new mutable specialization constants feature. See the [materials documentation](https://google.github.io/filament/Materials.html) for details. [⚠️ **New Material Version**]
## v1.61.1
@@ -60,6 +60,14 @@ static void setParameter(JNIEnv* env, jlong nativeMaterialInstance, jstring name
env->ReleaseStringUTFChars(name_, name);
}
template<typename T>
static void setConstant(JNIEnv* env, jlong nativeMaterialInstance, jstring name_, T v) {
MaterialInstance* instance = (MaterialInstance*) nativeMaterialInstance;
const char *name = env->GetStringUTFChars(name_, 0);
instance->setConstant(name, v);
env->ReleaseStringUTFChars(name_, name);
}
extern "C"
JNIEXPORT void JNICALL
Java_com_google_android_filament_MaterialInstance_nSetParameterBool(JNIEnv *env, jclass,
@@ -264,6 +272,13 @@ Java_com_google_android_filament_MaterialInstance_nSetParameterTexture(
env->ReleaseStringUTFChars(name_, name);
}
extern "C"
JNIEXPORT void JNICALL
Java_com_google_android_filament_MaterialInstance_nSetConstantBool(JNIEnv *env, jclass,
jlong nativeMaterialInstance, jstring name_, jboolean x) {
setConstant(env, nativeMaterialInstance, name_, bool(x));
}
extern "C"
JNIEXPORT void JNICALL
Java_com_google_android_filament_MaterialInstance_nSetScissor(
@@ -402,6 +402,16 @@ public class MaterialInstance {
nSetParameterFloat4(getNativeObject(), name, color[0], color[1], color[2], color[3]);
}
/**
* Sets the value of a bool constant.
*
* @param name the name of the material constant
* @param x the value of the material constant
*/
public void setConstant(@NonNull String name, boolean x) {
nSetConstantBool(getNativeObject(), name, x);
}
/**
* Set-up a custom scissor rectangle; by default it is disabled.
*
@@ -921,6 +931,9 @@ public class MaterialInstance {
@NonNull String name, int element, @NonNull @Size(min = 1) float[] v,
@IntRange(from = 0) int offset, @IntRange(from = 1) int count);
private static native void nSetConstantBool(long nativeMaterialInstance,
@NonNull String name, boolean x);
private static native void nSetParameterTexture(long nativeMaterialInstance,
@NonNull String name, long nativeTexture, long sampler);
+1 -1
View File
@@ -1,5 +1,5 @@
GROUP=com.google.android.filament
VERSION_NAME=1.62.0
VERSION_NAME=1.61.1
POM_DESCRIPTION=Real-time physically based rendering engine for Android.
Binary file not shown.
+1 -2
View File
@@ -1,7 +1,6 @@
#Wed Nov 17 10:40:18 PST 2021
distributionBase=GRADLE_USER_HOME
distributionPath=wrapper/dists
distributionUrl=https\://services.gradle.org/distributions/gradle-8.8-bin.zip
networkTimeout=10000
validateDistributionUrl=true
zipStoreBase=GRADLE_USER_HOME
zipStorePath=wrapper/dists
+109 -173
View File
@@ -1,7 +1,7 @@
#!/bin/sh
#!/usr/bin/env sh
#
# Copyright © 2015-2021 the original authors.
# Copyright 2015 the original author or authors.
#
# Licensed under the Apache License, Version 2.0 (the "License");
# you may not use this file except in compliance with the License.
@@ -17,99 +17,67 @@
#
##############################################################################
#
# Gradle start up script for POSIX generated by Gradle.
#
# Important for running:
#
# (1) You need a POSIX-compliant shell to run this script. If your /bin/sh is
# noncompliant, but you have some other compliant shell such as ksh or
# bash, then to run this script, type that shell name before the whole
# command line, like:
#
# ksh Gradle
#
# Busybox and similar reduced shells will NOT work, because this script
# requires all of these POSIX shell features:
# * functions;
# * expansions «$var», «${var}», «${var:-default}», «${var+SET}»,
# «${var#prefix}», «${var%suffix}», and «$( cmd )»;
# * compound commands having a testable exit status, especially «case»;
# * various built-in commands including «command», «set», and «ulimit».
#
# Important for patching:
#
# (2) This script targets any POSIX shell, so it avoids extensions provided
# by Bash, Ksh, etc; in particular arrays are avoided.
#
# The "traditional" practice of packing multiple parameters into a
# space-separated string is a well documented source of bugs and security
# problems, so this is (mostly) avoided, by progressively accumulating
# options in "$@", and eventually passing that to Java.
#
# Where the inherited environment variables (DEFAULT_JVM_OPTS, JAVA_OPTS,
# and GRADLE_OPTS) rely on word-splitting, this is performed explicitly;
# see the in-line comments for details.
#
# There are tweaks for specific operating systems such as AIX, CygWin,
# Darwin, MinGW, and NonStop.
#
# (3) This script is generated from the Groovy template
# https://github.com/gradle/gradle/blob/HEAD/platforms/jvm/plugins-application/src/main/resources/org/gradle/api/internal/plugins/unixStartScript.txt
# within the Gradle project.
#
# You can find Gradle at https://github.com/gradle/gradle/.
#
##
## Gradle start up script for UN*X
##
##############################################################################
# Attempt to set APP_HOME
# Resolve links: $0 may be a link
app_path=$0
# Need this for daisy-chained symlinks.
while
APP_HOME=${app_path%"${app_path##*/}"} # leaves a trailing /; empty if no leading path
[ -h "$app_path" ]
do
ls=$( ls -ld "$app_path" )
link=${ls#*' -> '}
case $link in #(
/*) app_path=$link ;; #(
*) app_path=$APP_HOME$link ;;
esac
PRG="$0"
# Need this for relative symlinks.
while [ -h "$PRG" ] ; do
ls=`ls -ld "$PRG"`
link=`expr "$ls" : '.*-> \(.*\)$'`
if expr "$link" : '/.*' > /dev/null; then
PRG="$link"
else
PRG=`dirname "$PRG"`"/$link"
fi
done
SAVED="`pwd`"
cd "`dirname \"$PRG\"`/" >/dev/null
APP_HOME="`pwd -P`"
cd "$SAVED" >/dev/null
# This is normally unused
# shellcheck disable=SC2034
APP_BASE_NAME=${0##*/}
# Discard cd standard output in case $CDPATH is set (https://github.com/gradle/gradle/issues/25036)
APP_HOME=$( cd "${APP_HOME:-./}" > /dev/null && pwd -P ) || exit
APP_NAME="Gradle"
APP_BASE_NAME=`basename "$0"`
# Add default JVM options here. You can also use JAVA_OPTS and GRADLE_OPTS to pass JVM options to this script.
DEFAULT_JVM_OPTS='"-Xmx64m" "-Xms64m"'
# Use the maximum available, or set MAX_FD != -1 to use that value.
MAX_FD=maximum
MAX_FD="maximum"
warn () {
echo "$*"
} >&2
}
die () {
echo
echo "$*"
echo
exit 1
} >&2
}
# OS specific support (must be 'true' or 'false').
cygwin=false
msys=false
darwin=false
nonstop=false
case "$( uname )" in #(
CYGWIN* ) cygwin=true ;; #(
Darwin* ) darwin=true ;; #(
MSYS* | MINGW* ) msys=true ;; #(
NONSTOP* ) nonstop=true ;;
case "`uname`" in
CYGWIN* )
cygwin=true
;;
Darwin* )
darwin=true
;;
MINGW* )
msys=true
;;
NONSTOP* )
nonstop=true
;;
esac
CLASSPATH=$APP_HOME/gradle/wrapper/gradle-wrapper.jar
@@ -119,9 +87,9 @@ CLASSPATH=$APP_HOME/gradle/wrapper/gradle-wrapper.jar
if [ -n "$JAVA_HOME" ] ; then
if [ -x "$JAVA_HOME/jre/sh/java" ] ; then
# IBM's JDK on AIX uses strange locations for the executables
JAVACMD=$JAVA_HOME/jre/sh/java
JAVACMD="$JAVA_HOME/jre/sh/java"
else
JAVACMD=$JAVA_HOME/bin/java
JAVACMD="$JAVA_HOME/bin/java"
fi
if [ ! -x "$JAVACMD" ] ; then
die "ERROR: JAVA_HOME is set to an invalid directory: $JAVA_HOME
@@ -130,120 +98,88 @@ Please set the JAVA_HOME variable in your environment to match the
location of your Java installation."
fi
else
JAVACMD=java
if ! command -v java >/dev/null 2>&1
then
die "ERROR: JAVA_HOME is not set and no 'java' command could be found in your PATH.
JAVACMD="java"
which java >/dev/null 2>&1 || die "ERROR: JAVA_HOME is not set and no 'java' command could be found in your PATH.
Please set the JAVA_HOME variable in your environment to match the
location of your Java installation."
fi
fi
# Increase the maximum file descriptors if we can.
if ! "$cygwin" && ! "$darwin" && ! "$nonstop" ; then
case $MAX_FD in #(
max*)
# In POSIX sh, ulimit -H is undefined. That's why the result is checked to see if it worked.
# shellcheck disable=SC2039,SC3045
MAX_FD=$( ulimit -H -n ) ||
warn "Could not query maximum file descriptor limit"
esac
case $MAX_FD in #(
'' | soft) :;; #(
*)
# In POSIX sh, ulimit -n is undefined. That's why the result is checked to see if it worked.
# shellcheck disable=SC2039,SC3045
ulimit -n "$MAX_FD" ||
warn "Could not set maximum file descriptor limit to $MAX_FD"
esac
if [ "$cygwin" = "false" -a "$darwin" = "false" -a "$nonstop" = "false" ] ; then
MAX_FD_LIMIT=`ulimit -H -n`
if [ $? -eq 0 ] ; then
if [ "$MAX_FD" = "maximum" -o "$MAX_FD" = "max" ] ; then
MAX_FD="$MAX_FD_LIMIT"
fi
ulimit -n $MAX_FD
if [ $? -ne 0 ] ; then
warn "Could not set maximum file descriptor limit: $MAX_FD"
fi
else
warn "Could not query maximum file descriptor limit: $MAX_FD_LIMIT"
fi
fi
# Collect all arguments for the java command, stacking in reverse order:
# * args from the command line
# * the main class name
# * -classpath
# * -D...appname settings
# * --module-path (only if needed)
# * DEFAULT_JVM_OPTS, JAVA_OPTS, and GRADLE_OPTS environment variables.
# For Darwin, add options to specify how the application appears in the dock
if $darwin; then
GRADLE_OPTS="$GRADLE_OPTS \"-Xdock:name=$APP_NAME\" \"-Xdock:icon=$APP_HOME/media/gradle.icns\""
fi
# For Cygwin or MSYS, switch paths to Windows format before running java
if "$cygwin" || "$msys" ; then
APP_HOME=$( cygpath --path --mixed "$APP_HOME" )
CLASSPATH=$( cygpath --path --mixed "$CLASSPATH" )
if [ "$cygwin" = "true" -o "$msys" = "true" ] ; then
APP_HOME=`cygpath --path --mixed "$APP_HOME"`
CLASSPATH=`cygpath --path --mixed "$CLASSPATH"`
JAVACMD=$( cygpath --unix "$JAVACMD" )
JAVACMD=`cygpath --unix "$JAVACMD"`
# Now convert the arguments - kludge to limit ourselves to /bin/sh
for arg do
if
case $arg in #(
-*) false ;; # don't mess with options #(
/?*) t=${arg#/} t=/${t%%/*} # looks like a POSIX filepath
[ -e "$t" ] ;; #(
*) false ;;
esac
then
arg=$( cygpath --path --ignore --mixed "$arg" )
fi
# Roll the args list around exactly as many times as the number of
# args, so each arg winds up back in the position where it started, but
# possibly modified.
#
# NB: a `for` loop captures its iteration list before it begins, so
# changing the positional parameters here affects neither the number of
# iterations, nor the values presented in `arg`.
shift # remove old arg
set -- "$@" "$arg" # push replacement arg
# We build the pattern for arguments to be converted via cygpath
ROOTDIRSRAW=`find -L / -maxdepth 1 -mindepth 1 -type d 2>/dev/null`
SEP=""
for dir in $ROOTDIRSRAW ; do
ROOTDIRS="$ROOTDIRS$SEP$dir"
SEP="|"
done
OURCYGPATTERN="(^($ROOTDIRS))"
# Add a user-defined pattern to the cygpath arguments
if [ "$GRADLE_CYGPATTERN" != "" ] ; then
OURCYGPATTERN="$OURCYGPATTERN|($GRADLE_CYGPATTERN)"
fi
# Now convert the arguments - kludge to limit ourselves to /bin/sh
i=0
for arg in "$@" ; do
CHECK=`echo "$arg"|egrep -c "$OURCYGPATTERN" -`
CHECK2=`echo "$arg"|egrep -c "^-"` ### Determine if an option
if [ $CHECK -ne 0 ] && [ $CHECK2 -eq 0 ] ; then ### Added a condition
eval `echo args$i`=`cygpath --path --ignore --mixed "$arg"`
else
eval `echo args$i`="\"$arg\""
fi
i=`expr $i + 1`
done
case $i in
0) set -- ;;
1) set -- "$args0" ;;
2) set -- "$args0" "$args1" ;;
3) set -- "$args0" "$args1" "$args2" ;;
4) set -- "$args0" "$args1" "$args2" "$args3" ;;
5) set -- "$args0" "$args1" "$args2" "$args3" "$args4" ;;
6) set -- "$args0" "$args1" "$args2" "$args3" "$args4" "$args5" ;;
7) set -- "$args0" "$args1" "$args2" "$args3" "$args4" "$args5" "$args6" ;;
8) set -- "$args0" "$args1" "$args2" "$args3" "$args4" "$args5" "$args6" "$args7" ;;
9) set -- "$args0" "$args1" "$args2" "$args3" "$args4" "$args5" "$args6" "$args7" "$args8" ;;
esac
fi
# Escape application args
save () {
for i do printf %s\\n "$i" | sed "s/'/'\\\\''/g;1s/^/'/;\$s/\$/' \\\\/" ; done
echo " "
}
APP_ARGS=`save "$@"`
# Add default JVM options here. You can also use JAVA_OPTS and GRADLE_OPTS to pass JVM options to this script.
DEFAULT_JVM_OPTS='"-Xmx64m" "-Xms64m"'
# Collect all arguments for the java command:
# * DEFAULT_JVM_OPTS, JAVA_OPTS, JAVA_OPTS, and optsEnvironmentVar are not allowed to contain shell fragments,
# and any embedded shellness will be escaped.
# * For example: A user cannot expect ${Hostname} to be expanded, as it is an environment variable and will be
# treated as '${Hostname}' itself on the command line.
set -- \
"-Dorg.gradle.appname=$APP_BASE_NAME" \
-classpath "$CLASSPATH" \
org.gradle.wrapper.GradleWrapperMain \
"$@"
# Stop when "xargs" is not available.
if ! command -v xargs >/dev/null 2>&1
then
die "xargs is not available"
fi
# Use "xargs" to parse quoted args.
#
# With -n1 it outputs one arg per line, with the quotes and backslashes removed.
#
# In Bash we could simply go:
#
# readarray ARGS < <( xargs -n1 <<<"$var" ) &&
# set -- "${ARGS[@]}" "$@"
#
# but POSIX shell has neither arrays nor command substitution, so instead we
# post-process each arg (as a line of input to sed) to backslash-escape any
# character that might be a shell metacharacter, then use eval to reverse
# that process (while maintaining the separation between arguments), and wrap
# the whole thing up as a single "set" statement.
#
# This will of course break if any of these variables contains a newline or
# an unmatched quote.
#
eval "set -- $(
printf '%s\n' "$DEFAULT_JVM_OPTS $JAVA_OPTS $GRADLE_OPTS" |
xargs -n1 |
sed ' s~[^-[:alnum:]+,./:=@_]~\\&~g; ' |
tr '\n' ' '
)" '"$@"'
# Collect all arguments for the java command, following the shell quoting and substitution rules
eval set -- $DEFAULT_JVM_OPTS $JAVA_OPTS $GRADLE_OPTS "\"-Dorg.gradle.appname=$APP_BASE_NAME\"" -classpath "\"$CLASSPATH\"" org.gradle.wrapper.GradleWrapperMain "$APP_ARGS"
exec "$JAVACMD" "$@"
+16 -19
View File
@@ -14,7 +14,7 @@
@rem limitations under the License.
@rem
@if "%DEBUG%"=="" @echo off
@if "%DEBUG%" == "" @echo off
@rem ##########################################################################
@rem
@rem Gradle startup script for Windows
@@ -25,8 +25,7 @@
if "%OS%"=="Windows_NT" setlocal
set DIRNAME=%~dp0
if "%DIRNAME%"=="" set DIRNAME=.
@rem This is normally unused
if "%DIRNAME%" == "" set DIRNAME=.
set APP_BASE_NAME=%~n0
set APP_HOME=%DIRNAME%
@@ -41,13 +40,13 @@ if defined JAVA_HOME goto findJavaFromJavaHome
set JAVA_EXE=java.exe
%JAVA_EXE% -version >NUL 2>&1
if %ERRORLEVEL% equ 0 goto execute
if "%ERRORLEVEL%" == "0" goto execute
echo. 1>&2
echo ERROR: JAVA_HOME is not set and no 'java' command could be found in your PATH. 1>&2
echo. 1>&2
echo Please set the JAVA_HOME variable in your environment to match the 1>&2
echo location of your Java installation. 1>&2
echo.
echo ERROR: JAVA_HOME is not set and no 'java' command could be found in your PATH.
echo.
echo Please set the JAVA_HOME variable in your environment to match the
echo location of your Java installation.
goto fail
@@ -57,11 +56,11 @@ set JAVA_EXE=%JAVA_HOME%/bin/java.exe
if exist "%JAVA_EXE%" goto execute
echo. 1>&2
echo ERROR: JAVA_HOME is set to an invalid directory: %JAVA_HOME% 1>&2
echo. 1>&2
echo Please set the JAVA_HOME variable in your environment to match the 1>&2
echo location of your Java installation. 1>&2
echo.
echo ERROR: JAVA_HOME is set to an invalid directory: %JAVA_HOME%
echo.
echo Please set the JAVA_HOME variable in your environment to match the
echo location of your Java installation.
goto fail
@@ -76,15 +75,13 @@ set CLASSPATH=%APP_HOME%\gradle\wrapper\gradle-wrapper.jar
:end
@rem End local scope for the variables with windows NT shell
if %ERRORLEVEL% equ 0 goto mainEnd
if "%ERRORLEVEL%"=="0" goto mainEnd
:fail
rem Set variable GRADLE_EXIT_CONSOLE if you need the _script_ return code instead of
rem the _cmd.exe /c_ return code!
set EXIT_CODE=%ERRORLEVEL%
if %EXIT_CODE% equ 0 set EXIT_CODE=1
if not ""=="%GRADLE_EXIT_CONSOLE%" exit %EXIT_CODE%
exit /b %EXIT_CODE%
if not "" == "%GRADLE_EXIT_CONSOLE%" exit 1
exit /b 1
:mainEnd
if "%OS%"=="Windows_NT" endlocal
+6 -7
View File
@@ -204,10 +204,9 @@
"license": "Apache-2.0"
},
"node_modules/brace-expansion": {
"version": "1.1.12",
"resolved": "https://registry.npmjs.org/brace-expansion/-/brace-expansion-1.1.12.tgz",
"integrity": "sha512-9T9UjW3r0UW5c1Q7GTwllptXwhvYmEzFhzMfZ9H7FQWt+uZePjZPjBP/W1ZEyZ1twGWom5/56TF4lPcqjnDHcg==",
"license": "MIT",
"version": "1.1.11",
"resolved": "https://registry.npmjs.org/brace-expansion/-/brace-expansion-1.1.11.tgz",
"integrity": "sha512-iCuPHDFgrHX7H2vEI/5xpz07zSHB00TpugqhmYtVmMO6518mCuRMoOYFldEBl0g187ufozdaHgWKcYFb61qGiA==",
"dependencies": {
"balanced-match": "^1.0.0",
"concat-map": "0.0.1"
@@ -410,9 +409,9 @@
"integrity": "sha512-Nik3Sc0ncrMK4UUdXQmAnRtzmNQTAAXmXIopizwZ1W1t8QmfJj+zL4OA2I7XPTPW5z5TDqv4hRo/JzouDJnX3A=="
},
"brace-expansion": {
"version": "1.1.12",
"resolved": "https://registry.npmjs.org/brace-expansion/-/brace-expansion-1.1.12.tgz",
"integrity": "sha512-9T9UjW3r0UW5c1Q7GTwllptXwhvYmEzFhzMfZ9H7FQWt+uZePjZPjBP/W1ZEyZ1twGWom5/56TF4lPcqjnDHcg==",
"version": "1.1.11",
"resolved": "https://registry.npmjs.org/brace-expansion/-/brace-expansion-1.1.11.tgz",
"integrity": "sha512-iCuPHDFgrHX7H2vEI/5xpz07zSHB00TpugqhmYtVmMO6518mCuRMoOYFldEBl0g187ufozdaHgWKcYFb61qGiA==",
"requires": {
"balanced-match": "^1.0.0",
"concat-map": "0.0.1"
+2 -5
View File
@@ -1027,11 +1027,8 @@ samplerCubemap | Cubemap texture
[Table [materialParamsTypes]: Material parameter types]
Samplers
: Sampler types can have the following fields:
- `format` : which can be either `int` or `float` (defaults to `float`).
- `multisample` : a boolean to indicate whether the sampler is meant for multisampling (defaults to `false`)
- `unfilterable` : a boolean to indicate whether the sampling is not filtered (defaults to `false`)
: Sampler types can also specify a `format` which can be either `int` or `float` (defaults to
`float`).
Arrays
: A parameter can define an array of values by appending `[size]` after the type name, where
+31 -24
View File
@@ -211,7 +211,7 @@ when using textures.
This property can dramatically change the appearance of a surface. Non-metallic surfaces have
chromatic diffuse reflection and achromatic specular reflection (reflected light does not change
color). Metallic surfaces do not have any diffuse reflection and chromatic specular reflection
(reflected light takes on the color of the surfaced as defined by `baseColor`).
(reflected light takes on the color of the surfaced as defined by `baseColor`).
The effect of `metallic` is shown in figure [metallicProperty] (click on the image to see a
larger version).
@@ -247,7 +247,7 @@ The effect of `roughness` on metallic surfaces is shown in figure [roughnessCond
When refraction through an object is enabled (using a `refractonType` of `thin` or `solid`), the
`roughness` property will also affect the refractions, as shown in figure
[roughnessRefractionProperty] (click on the image to see a larger version).
![Figure [roughnessRefractionProperty]: Refractive sphere with `roughness` varying from 0.0
(left) to 1.0 (right)](images/materials/refraction_roughness.png)
@@ -306,7 +306,7 @@ The sheen color controls the color appearance and strength of an optional sheen
base layer described by the properties above. The sheen layer always sits below the clear coat layer
if such a layer is present.
The sheen layer can be used to represent cloth and fabric materials. Please refer to
The sheen layer can be used to represent cloth and fabric materials. Please refer to
section [Cloth model] for more information about cloth and fabric materials.
The effect of `sheenColor` is shown in figure [materialSheenColor]
@@ -519,14 +519,14 @@ light to bend further away from the initial path.
Table [commonMatIOR] describes acceptable refractive indices for various types of materials.
Material | IOR
Material | IOR
--------------------------:|:-----------------
Air | 1.0
Water | 1.33
Common liquids | 1.33 to 1.5
Common gemstones | 1.58 to 2.33
Plastics, glass | 1.5 to 1.58
Other dielectric materials | 1.33 to 1.58
Water | 1.33
Common liquids | 1.33 to 1.5
Common gemstones | 1.58 to 2.33
Plastics, glass | 1.5 to 1.58
Other dielectric materials | 1.33 to 1.58
[Table [commonMatIOR]: Index of refraction of common materials]
The appearance of a refractive material will greatly depend on the `refractionType` and
@@ -1091,30 +1091,37 @@ Value
`bool` or `number`, depending on the `type` of the constant. The type must be one of the types
described in table [materialConstantsTypes].
Type | Description | Default
:----------------------|:-----------------------------------------|:------------------
int | A signed, 32 bit GLSL int | 0
float | A single-precision GLSL float | 0.0
bool | A GLSL bool | false
Constants may also be specified as mutable by setting the `mutable` property to `true`. Only
`bool` spec constants may be specified as mutable.
| Type | Description | Default | May be mutable? |
|:------|:------------------------------|:--------|-----------------|
| int | A signed, 32 bit GLSL int | | no |
| float | A single-precision GLSL float | 0.0 | no |
| bool | A GLSL bool | false | yes |
[Table [materialConstantsTypes]: Material constants types]
Description
: Lists the constant parameters accepted by your material. These constants can be set, or
"specialized", at runtime when loading a material package. Multiple materials can be loaded from
the same material package with differing constant parameter specializations. Once a material is
loaded from a material package, its constant parameters cannot be changed. Compared to regular
parameters, constant parameters allow the compiler to generate more efficient code. Access
constant parameters from the shader by prefixing the name with `materialConstant_`. For example,
a constant parameter named `myConstant` is accessed in the shader as
`materialConstant_myConstant`. If a constant parameter is not set at runtime, the default is
used.
the same material package with differing constant parameter specializations. If a constant
parameter is specialized as mutable, it may be changed at any time on a per-instance basis via
`MaterialInstance::setConstant(name, value)`. Otherwise, once a material is loaded from a
material package, its constant parameters cannot be changed.
Compared to regular parameters, constant parameters allow the compiler to generate more
efficient code. Access constant parameters from the shader by prefixing the name with
`materialConstant_`. For example, a constant parameter named `myConstant` is accessed in the
shader as `materialConstant_myConstant`. If a constant parameter is not set at runtime, the
default is used.
~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~ JSON
material {
constants : [
{
name : overrideAlpha,
type : bool
type : bool,
mutable : true,
},
{
name : customAlpha,
@@ -2553,7 +2560,7 @@ standard skybox material. It produces a list of 2 parameters, named `showSun` an
respectively a boolean and a cubemap texture.
```text
$ matc --reflect parameters filament/src/materials/skybox.mat
$ matc --reflect parameters filament/src/materials/skybox.mat
{
"parameters": [
{
@@ -2569,7 +2576,7 @@ $ matc --reflect parameters filament/src/materials/skybox.mat
}
]
}
```
```
### --variant-filter
+1 -13
View File
@@ -257,7 +257,6 @@ if (FILAMENT_SUPPORTS_WEBGPU)
list(APPEND SRCS
include/backend/platforms/WebGPUPlatform.h
src/webgpu/platform/WebGPUPlatform.cpp
src/webgpu/SpdMipmapGenerator/SpdMipmapGenerator.cpp
src/webgpu/WebGPUBufferBase.cpp
src/webgpu/WebGPUBufferBase.h
src/webgpu/WebGPUBufferObject.cpp
@@ -269,8 +268,6 @@ if (FILAMENT_SUPPORTS_WEBGPU)
src/webgpu/WebGPUDescriptorSetLayout.h
src/webgpu/WebGPUDriver.cpp
src/webgpu/WebGPUDriver.h
src/webgpu/WebGPUFence.cpp
src/webgpu/WebGPUFence.h
src/webgpu/WebGPUIndexBuffer.cpp
src/webgpu/WebGPUIndexBuffer.h
src/webgpu/WebGPUPipelineCreation.cpp
@@ -505,20 +502,11 @@ if (FILAMENT_SUPPORTS_WEBGPU)
target_compile_definitions(${TARGET} PRIVATE $<$<BOOL:${FILAMENT_WEBGPU_IMMEDIATE_ERROR_HANDLING}>:FILAMENT_WEBGPU_IMMEDIATE_ERROR_HANDLING>)
endif()
# ==================================================================================================
# Linker flags
# ==================================================================================================
target_link_libraries(${TARGET} PRIVATE
${OSMESA_LINKER_FLAGS}
$<$<AND:$<PLATFORM_ID:Linux>,$<CONFIG:Release>>:${LINUX_LINKER_OPTIMIZATION_FLAGS}>
)
if (LINUX AND FILAMENT_SUPPORTS_WEBGPU)
# Unclear why this needs to be explicit while other linux libs do not, but it is necessary for
# WebGPU to build on linux.
target_link_libraries(${TARGET} PRIVATE -lxcb)
endif()
# ==================================================================================================
# Installation
# ==================================================================================================
+112 -6
View File
@@ -474,12 +474,7 @@ constexpr std::string_view to_string(DescriptorType type) noexcept {
enum class DescriptorFlags : uint8_t {
NONE = 0x00,
// Indicate a UNIFORM_BUFFER will have dynamic offsets.
DYNAMIC_OFFSET = 0x01,
// To indicate a texture/sampler type should be unfiltered.
UNFILTERABLE = 0x02,
DYNAMIC_OFFSET = 0x01
};
using descriptor_set_t = uint8_t;
@@ -500,12 +495,18 @@ struct DescriptorSetLayoutBinding {
DescriptorFlags flags = DescriptorFlags::NONE;
uint16_t count = 0;
// TODO: uncomment when needed. Note that this class is used as hash key. We need to ensure
// no uninitialized padding bytes.
// uint8_t externalSamplerDataIndex = EXTERNAL_SAMPLER_DATA_INDEX_UNUSED;
friend bool operator==(DescriptorSetLayoutBinding const& lhs,
DescriptorSetLayoutBinding const& rhs) noexcept {
return lhs.type == rhs.type &&
lhs.flags == rhs.flags &&
lhs.count == rhs.count &&
lhs.stageFlags == rhs.stageFlags;
// lhs.stageFlags == rhs.stageFlags &&
// lhs.externalSamplerDataIndex == rhs.externalSamplerDataIndex;
}
};
@@ -1253,6 +1254,26 @@ enum class SamplerCompareFunc : uint8_t {
N //!< Never. The depth / stencil test always fails.
};
//! this API is copied from (and only applies to) the Vulkan spec.
//! These specify YUV to RGB conversions.
enum class SamplerYcbcrModelConversion : uint8_t {
RGB_IDENTITY = 0,
YCBCR_IDENTITY = 1,
YCBCR_709 = 2,
YCBCR_601 = 3,
YCBCR_2020 = 4,
};
enum class SamplerYcbcrRange : uint8_t {
ITU_FULL = 0,
ITU_NARROW = 1,
};
enum class ChromaLocation : uint8_t {
COSITED_EVEN = 0,
MIDPOINT = 1,
};
//! Sampler parameters
struct SamplerParams { // NOLINT
SamplerMagFilter filterMag : 1; //!< magnification filter (NEAREST)
@@ -1321,9 +1342,94 @@ static_assert(sizeof(SamplerParams) == 4);
static_assert(sizeof(SamplerParams) <= sizeof(uint64_t),
"SamplerParams must be no more than 64 bits");
//! Sampler parameters
struct SamplerYcbcrConversion {// NOLINT
SamplerYcbcrModelConversion ycbcrModel : 4;
TextureSwizzle r : 4;
TextureSwizzle g : 4;
TextureSwizzle b : 4;
TextureSwizzle a : 4;
SamplerYcbcrRange ycbcrRange : 1;
ChromaLocation xChromaOffset : 1;
ChromaLocation yChromaOffset : 1;
SamplerMagFilter chromaFilter : 1;
uint8_t padding;
struct Hasher {
size_t operator()(const SamplerYcbcrConversion p) const noexcept {
// we don't use std::hash<> here, so we don't have to include <functional>
return *reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&p));
}
};
struct EqualTo {
bool operator()(SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs) const noexcept {
assert_invariant(lhs.padding == 0);
auto* pLhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&lhs));
auto* pRhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&rhs));
return *pLhs == *pRhs;
}
};
struct LessThan {
bool operator()(SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs) const noexcept {
assert_invariant(lhs.padding == 0);
auto* pLhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&lhs));
auto* pRhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&rhs));
return *pLhs < *pRhs;
}
};
private:
friend bool operator == (SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs)
noexcept {
return SamplerYcbcrConversion::EqualTo{}(lhs, rhs);
}
friend bool operator != (SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs)
noexcept {
return !SamplerYcbcrConversion::EqualTo{}(lhs, rhs);
}
friend bool operator < (SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs)
noexcept {
return SamplerYcbcrConversion::LessThan{}(lhs, rhs);
}
};
static_assert(sizeof(SamplerYcbcrConversion) == 4);
static_assert(sizeof(SamplerYcbcrConversion) <= sizeof(uint64_t),
"SamplerYcbcrConversion must be no more than 64 bits");
struct ExternalSamplerDatum {
ExternalSamplerDatum(SamplerYcbcrConversion ycbcr, SamplerParams spm, uint32_t extFmt)
: YcbcrConversion(ycbcr),
samplerParams(spm),
externalFormat(extFmt) {}
bool operator==(ExternalSamplerDatum const& rhs) const {
return (YcbcrConversion == rhs.YcbcrConversion && samplerParams == rhs.samplerParams &&
externalFormat == rhs.externalFormat);
}
struct EqualTo {
bool operator()(const ExternalSamplerDatum& lhs,
const ExternalSamplerDatum& rhs) const noexcept {
return (lhs.YcbcrConversion == rhs.YcbcrConversion &&
lhs.samplerParams == rhs.samplerParams &&
lhs.externalFormat == rhs.externalFormat);
}
};
SamplerYcbcrConversion YcbcrConversion;
SamplerParams samplerParams;
uint32_t externalFormat;
};
// No implicit padding allowed due to it being a hash key.
static_assert(sizeof(ExternalSamplerDatum) == 12);
struct DescriptorSetLayout {
std::variant<utils::StaticString, utils::CString, std::monostate> label;
utils::FixedCapacityVector<DescriptorSetLayoutBinding> bindings;
// TODO: uncomment when needed
// utils::FixedCapacityVector<ExternalSamplerDatum> externalSamplerData;
};
//! blending equation function
@@ -107,36 +107,6 @@ public:
MULTIVIEW,
};
/**
* This controls the priority level for GPU work scheduling, which helps prioritize the
* submitted GPU work and enables preemption.
*/
enum class GpuContextPriority : uint8_t {
/**
* Backend default GPU context priority (typically MEDIUM)
*/
DEFAULT,
/**
* For non-interactive, deferrable workloads. This should not interfere with standard
* applications.
*/
LOW,
/**
* The default priority level for standard applications.
*/
MEDIUM,
/**
* For high-priority, latency-sensitive workloads that are more important than standard
* applications.
*/
HIGH,
/**
* The highest priority, intended for system-critical, real-time applications where missing
* deadlines is unacceptable (e.g., VR/AR compositors or other system-critical tasks).
*/
REALTIME,
};
struct DriverConfig {
/**
* Size of handle arena in bytes. Setting to 0 indicates default value is to be used.
@@ -179,20 +149,6 @@ public:
* - PlatformEGLAndroid
*/
bool assertNativeWindowIsValid = false;
/**
* The action to take if a Drawable cannot be acquired. If true, the
* frame is aborted instead of panic. This is only supported for:
* - PlatformMetal
*/
bool metalDisablePanicOnDrawableFailure = false;
/**
* GPU context priority level. Controls GPU work scheduling and preemption.
* This is only supported for:
* - PlatformEGL
*/
GpuContextPriority gpuContextPriority = GpuContextPriority::DEFAULT;
};
Platform() noexcept;
+4 -1
View File
@@ -20,6 +20,7 @@
#include <utils/CString.h>
#include <utils/FixedCapacityVector.h>
#include <utils/Invocable.h>
#include <utils/bitset.h>
#include <backend/DriverEnums.h>
@@ -66,6 +67,7 @@ public:
using DescriptorBindingsInfo = utils::FixedCapacityVector<Descriptor>;
using DescriptorSetInfo = std::array<DescriptorBindingsInfo, MAX_DESCRIPTOR_SET_COUNT>;
using SpecializationConstantsInfo = utils::FixedCapacityVector<SpecializationConstant>;
using MutableSpecConstantsInfo = utils::bitset8;
using ShaderBlob = utils::FixedCapacityVector<uint8_t>;
using ShaderSource = std::array<ShaderBlob, SHADER_TYPE_COUNT>;
@@ -102,7 +104,8 @@ public:
Program& descriptorBindings(backend::descriptor_set_t set,
DescriptorBindingsInfo descriptorBindings) noexcept;
Program& specializationConstants(SpecializationConstantsInfo specConstants) noexcept;
Program& specializationConstants(SpecializationConstantsInfo specConstants,
uint32_t firstMutableId, MutableSpecConstantsInfo mutableSpecConstants) noexcept;
struct PushConstant {
utils::CString name;
@@ -116,7 +116,7 @@ public:
* Destroys the object D at Handle<B> and construct a new D in its place
* e.g.:
* Handle<ConcreteTexture> h = allocateAndConstruct(w, h);
* ConcreteTexture* p = destroyAndConstruct(h, w, h);
* ConcreteTexture* p = reconstruct(h, w, h);
*/
template<typename D, typename B, typename ... ARGS>
std::enable_if_t<std::is_base_of_v<B, D>, D>*
+11 -1
View File
@@ -80,8 +80,18 @@ Program& Program::attributes(AttributesInfo attributes) noexcept {
return *this;
}
Program& Program::specializationConstants(SpecializationConstantsInfo specConstants) noexcept {
Program& Program::specializationConstants(SpecializationConstantsInfo specConstants,
uint32_t firstMutableId, MutableSpecConstantsInfo mutableSpecConstants) noexcept {
// String the two lists together.
mSpecializationConstants = std::move(specConstants);
uint32_t firstMutableIndex = specConstants.size();
mSpecializationConstants.reserve(specConstants.size() + mutableSpecConstants.size());
for (uint32_t i = 0; i < mutableSpecConstants.size(); i++) {
mSpecializationConstants[i + firstMutableIndex] = SpecializationConstant {
.id = i + firstMutableId,
.value = mutableSpecConstants[i],
};
}
return *this;
}
@@ -53,9 +53,6 @@ PlatformMetal::~PlatformMetal() noexcept {
}
Driver* PlatformMetal::createDriver(void* /*sharedContext*/, const Platform::DriverConfig& driverConfig) noexcept {
pImpl->mDrawableFailureBehavior = driverConfig.metalDisablePanicOnDrawableFailure
? DrawableFailureBehavior::ABORT_FRAME
: DrawableFailureBehavior::PANIC;
return MetalDriverFactory::create(this, driverConfig);
}
@@ -122,18 +122,12 @@ struct ShaderCompilerService::OpenGLProgramToken : ProgramToken {
}
std::optional<CallbackManager::Handle> handle{};
// Only valid when the blob functions are provided by users. The validity of this variable
// doesn't guarantee that the program was created from the cache blob.
BlobCacheKey key;
// Used for the `THREAD_POOL` mode.
mutable Mutex lock;
mutable Condition cond;
bool signaled = false;
// Indicate this program was created from the cache blob.
bool retrievedFromBlobCache = false;
};
ShaderCompilerService::OpenGLProgramToken::~OpenGLProgramToken() {
@@ -265,7 +259,6 @@ ShaderCompilerService::program_token_t ShaderCompilerService::createProgram(
// Try retrieving the cached program blob if available.
token->gl.program = mBlobCache.retrieve(&token->key, mDriver.mPlatform, program);
if (token->gl.program) {
token->retrievedFromBlobCache = true;
return token;
}
@@ -344,15 +337,12 @@ GLuint ShaderCompilerService::getProgram(program_token_t& token) {
assert_invariant(token);// This function should be called when the token is still alive.
// Finalize any pending shader compilation tasks only when the token was created without cache.
if (!token->retrievedFromBlobCache) {
if (token->compiler.mMode == Mode::THREAD_POOL) {
auto const job = token->compiler.mCompilerThreadPool.dequeue(token);
if (!job) {
// It's likely that the job was already completed. But it may be still being
// executed at this moment. Just try waiting for it to avoid a race.
token->wait();
}
if (token->compiler.mMode == Mode::THREAD_POOL) {
auto const job = token->compiler.mCompilerThreadPool.dequeue(token);
if (!job) {
// It's likely that the job was already completed. But it may be still being
// executed at this moment. Just try waiting for it to avoid a race.
token->wait();
}
}
@@ -209,42 +209,10 @@ Driver* PlatformEGL::createDriver(void* sharedContext, const DriverConfig& drive
}
#endif
// Configure GPU context priority level for scheduling and preemption
if (driverConfig.gpuContextPriority != Platform::GpuContextPriority::DEFAULT) {
if (extensions.has("EGL_IMG_context_priority")) {
EGLint priorityLevel = EGL_CONTEXT_PRIORITY_MEDIUM_IMG;
const char* priorityName;
switch (driverConfig.gpuContextPriority) {
case Platform::GpuContextPriority::DEFAULT:
assert_invariant(false);
break;
case Platform::GpuContextPriority::LOW:
priorityLevel = EGL_CONTEXT_PRIORITY_LOW_IMG;
priorityName = "LOW";
break;
case Platform::GpuContextPriority::MEDIUM:
priorityLevel = EGL_CONTEXT_PRIORITY_MEDIUM_IMG;
priorityName = "MEDIUM";
break;
case Platform::GpuContextPriority::HIGH:
priorityLevel = EGL_CONTEXT_PRIORITY_HIGH_IMG;
priorityName = "HIGH";
break;
case Platform::GpuContextPriority::REALTIME:
priorityLevel = EGL_CONTEXT_PRIORITY_HIGH_IMG;
priorityName = "REALTIME(=HIGH)";
break;
}
contextAttribs[EGL_CONTEXT_PRIORITY_LEVEL_IMG] = priorityLevel;
LOG(INFO) << "EGL: Enabling GPU context priority: " << priorityName;
} else {
LOG(WARNING) << "EGL: GPU context priority requested but not supported";
}
}
// config use for creating the context
EGLConfig eglConfig = EGL_NO_CONFIG_KHR;
if (UTILS_UNLIKELY(!ext.egl.KHR_no_config_context)) {
// find a config we can use if we don't have "EGL_KHR_no_config_context" and that we can use
// for the dummy pbuffer surface.
@@ -15,10 +15,10 @@
*/
#include "VulkanBufferProxy.h"
#include "VulkanCommands.h"
#include "VulkanMemory.h"
#include "VulkanBufferCache.h"
#include "VulkanCommands.h"
#include "VulkanContext.h"
#include "VulkanMemory.h"
using namespace bluevk;
@@ -27,8 +27,7 @@ namespace filament::backend {
VulkanBufferProxy::VulkanBufferProxy(VmaAllocator allocator, VulkanStagePool& stagePool,
VulkanBufferCache& bufferCache, VulkanBufferUsage usage, uint32_t numBytes)
: mAllocator(allocator),
mStagePool(stagePool),
: mStagePool(stagePool),
mBufferCache(bufferCache),
mBuffer(mBufferCache.acquire(usage, numBytes)),
mUpdatedOffset(0),
@@ -36,45 +35,11 @@ VulkanBufferProxy::VulkanBufferProxy(VmaAllocator allocator, VulkanStagePool& st
void VulkanBufferProxy::loadFromCpu(VulkanCommandBuffer& commands, const void* cpuData,
uint32_t byteOffset, uint32_t numBytes) {
// The VulkanBuffer is available if the only object reference is hold by the
// `VulkanBufferProxy`. This means that the buffer is not currently referenced by a
// `VulkanCommandBuffer`.
bool const isAvailable = mBuffer->getCount() == 1;
if (isAvailable) {
// We are up to date with all the update operations, so no need to synchronize with previous
// updates.
mUpdatedBytes = 0;
mUpdatedOffset = 0;
}
// Keep track of the VulkanBuffer usage
commands.acquire(mBuffer);
// Check if we can just memcpy directly to the GPU memory.
// This is only allowed for UNIFORMS that are AVAILABLE and the memory is HOST_VISIBLE
// (supports memcpy from host). This works regardless if it's a full or partial update of the
// buffer.
bool const isMemcopyable = mBuffer->getGpuBuffer()->allocationInfo.pMappedData != nullptr;
bool const isUniform = getUsage() == VulkanBufferUsage::UNIFORM;
bool const useMemcpy =
isUniform && isMemcopyable && isAvailable && !FVK_FORCE_STAGING_FOR_BUFFER_UPDATES;
if (useMemcpy) {
char* dest = static_cast<char*>(mBuffer->getGpuBuffer()->allocationInfo.pMappedData) +
byteOffset;
memcpy(dest, cpuData, numBytes);
vmaFlushAllocation(mAllocator, mBuffer->getGpuBuffer()->vmaAllocation, byteOffset,
numBytes);
return;
}
// Note: this should be stored within the command buffer before going out of
// scope, so that the command buffer can manage its lifecycle.
fvkmemory::resource_ptr<VulkanStage::Segment> stage = mStagePool.acquireStage(numBytes);
assert_invariant(stage->memory());
commands.acquire(stage);
memcpy(stage->mapping(), cpuData, numBytes);
vmaFlushAllocation(mAllocator, stage->memory(), stage->offset(), numBytes);
stage->copy(0, cpuData, numBytes);
// If there was a previous update, then we need to make sure the following write is properly
// synced with the previous read.
@@ -160,4 +125,4 @@ VulkanBufferUsage VulkanBufferProxy::getUsage() const noexcept {
return mBuffer->getGpuBuffer()->usage;
}
} // namespace filament::backend
}// namespace filament::backend
@@ -40,7 +40,6 @@ public:
VulkanBufferUsage getUsage() const noexcept;
private:
VmaAllocator mAllocator;
VulkanStagePool& mStagePool;
VulkanBufferCache& mBufferCache;
@@ -49,6 +48,6 @@ private:
uint32_t mUpdatedBytes = 0;
};
} // namespace filament::backend
}// namespace filament::backend
#endif // TNT_FILAMENT_BACKEND_VULKANBUFFERPROXY_H
#endif// TNT_FILAMENT_BACKEND_VULKANBUFFERPROXY_H
@@ -95,7 +95,7 @@
#endif
#ifndef NDEBUG
#define FVK_DEBUG_FLAGS (FVK_DEBUG_PERFORMANCE | FVK_DEBUG_FORWARDED_FLAG)
#define FVK_DEBUG_FLAGS (FVK_DEBUG_PERFORMANCE | FVK_DEBUG_FORWARDED_FLAG | FVK_DEBUG_DEBUG_UTILS | FVK_DEBUG_GROUP_MARKERS | FVK_DEBUG_VALIDATION)
#else
#define FVK_DEBUG_FLAGS 0
#endif
@@ -214,8 +214,4 @@ constexpr static const int FVK_MAX_PIPELINE_AGE = FVK_MAX_COMMAND_BUFFERS;
// destroying any unused pipeline object.
static_assert(FVK_MAX_PIPELINE_AGE >= FVK_MAX_COMMAND_BUFFERS);
// Whether the buffer updates always use staging or not. Otherwise its allowed to also use memcpy
// for buffer updates. Default is false
constexpr static const bool FVK_FORCE_STAGING_FOR_BUFFER_UPDATES = false;
#endif
+2 -9
View File
@@ -394,13 +394,7 @@ void VulkanDriver::collectGarbage() {
void VulkanDriver::beginFrame(int64_t monotonic_clock_ns,
int64_t refreshIntervalNs, uint32_t frameId) {
FVK_PROFILE_MARKER(PROFILE_NAME_BEGINFRAME);
// Check if any command have finished and reset all its used resources. The resources
// wont be destroyed but their reference count will decreased if the command is already
// completed.
//
// This will let us check if any VulkanBuffer is currently in flight or not.
mCommands.gc();
// Do nothing.
if (mAppState.hasExternalSamplers()) {
mExternalImageManager.onBeginFrame();
@@ -1242,8 +1236,7 @@ void VulkanDriver::updateIndexBuffer(Handle<HwIndexBuffer> ibh, BufferDescriptor
scheduleDestroy(std::move(p));
}
void VulkanDriver::registerBufferObjectStreams(Handle<HwBufferObject> boh,
BufferObjectStreamDescriptor&& streams) {
void VulkanDriver::registerBufferObjectStreams(Handle<HwBufferObject> boh, BufferObjectStreamDescriptor&& streams) {
// Noop
}
+29 -42
View File
@@ -34,26 +34,27 @@ namespace {
static constexpr uint32_t MAX_EMPTY_STAGES_TO_RETAIN = 1;
constexpr uint32_t STAGE_SIZE = 1048576;
uint32_t alignValue(uint32_t valueToAlign, uint32_t alignment) {
if (alignment == 0) {
return valueToAlign;
}
uint32_t remainder = valueToAlign % alignment;
return remainder == 0 ? valueToAlign : valueToAlign + (alignment - remainder);
}
}// namespace
fvkmemory::resource_ptr<VulkanStage::Segment> VulkanStage::acquireSegment(
fvkmemory::ResourceManager* resManager, uint32_t segmentOffset, uint32_t numBytes) {
auto segment = fvkmemory::resource_ptr<Segment>::construct(resManager, this, numBytes,
segmentOffset, [this](uint32_t offset) { mSegments.erase(offset); });
mSegments.insert({ segmentOffset, segment.get() });
mCurrentOffset = segmentOffset + numBytes;
fvkmemory::ResourceManager* resManager, uint32_t numBytes) {
auto segment = fvkmemory::resource_ptr<Segment>::construct(resManager, mAllocator, this,
mCurrentOffset, [this](uint32_t offset) { mSegments.erase(offset); });
mSegments.insert({ mCurrentOffset, segment.get() });
// constexpr uint32_t BLOCK_SIZE = 16;
// uint32_t const additional = BLOCK_SIZE - (numBytes % BLOCK_SIZE);
// numBytes += (additional != BLOCK_SIZE ? additional : 0);
mCurrentOffset += numBytes;
return segment;
}
VkResult VulkanStage::Segment::copy(size_t dstOffset, void const* src, size_t writeSize) {
// uint8_t* mapped = ((uint8_t*) mapping()) + dstOffset;
// memcpy(mapped, src, writeSize);
// return vmaFlushAllocation(mAllocator, memory(), offset() + dstOffset, writeSize);
return vmaCopyMemoryToAllocation(mAllocator, src, memory(), offset() + dstOffset, writeSize);
}
VulkanStagePool::VulkanStagePool(VmaAllocator allocator, fvkmemory::ResourceManager* resManager,
VulkanCommands* commands, const VkPhysicalDeviceLimits* deviceLimits)
: mAllocator(allocator),
@@ -61,8 +62,7 @@ VulkanStagePool::VulkanStagePool(VmaAllocator allocator, fvkmemory::ResourceMana
mCommands(commands),
mDeviceLimits(deviceLimits) {}
fvkmemory::resource_ptr<VulkanStage::Segment> VulkanStagePool::acquireStage(uint32_t numBytes,
uint32_t alignment) {
fvkmemory::resource_ptr<VulkanStage::Segment> VulkanStagePool::acquireStage(uint32_t numBytes) {
// Apply alignment to the byte count to ensure that, when we later flush
// data written by the host, we only flush the atoms that we modified, and
// no adjacent atoms.
@@ -72,35 +72,17 @@ fvkmemory::resource_ptr<VulkanStage::Segment> VulkanStagePool::acquireStage(uint
// equal to the requested size.
auto iter = mStages.lower_bound(numBytes);
// If we have potential buffer candidates, loop through and try to find one
// that works with our alignment. Also, set the offset for the buffer.
// If we have no buffer, create a new one.
VulkanStage* pStage = nullptr;
uint32_t segmentOffset = 0;
while (iter != mStages.end()) {
segmentOffset = alignValue(iter->second->currentOffset(), alignment);
// Check for overflow + if there's space.
if (segmentOffset >= iter->second->currentOffset() &&
iter->second->capacity() - numBytes >= segmentOffset) {
// This buffer has enough space.
pStage = iter->second;
mStages.erase(iter);
break;
}
// This buffer wasn't suitable. Check the next larger one.
++iter;
}
if (!pStage) {
VulkanStage* pStage;
if (iter != mStages.end()) {
pStage = iter->second;
mStages.erase(iter);
} else {
pStage = allocateNewStage(std::max(numBytes, STAGE_SIZE));
}
// Note: this allocation updates `currentOffset` and `segments` within
// the parent stage. When destroyed, it will update `segments`.
fvkmemory::resource_ptr<VulkanStage::Segment> pSegment =
pStage->acquireSegment(mResManager, segmentOffset, numBytes);
fvkmemory::resource_ptr<VulkanStage::Segment> pSegment = pStage->acquireSegment(mResManager, numBytes);
// Update the stage's metadata, and reinsert it with the remaining segment
// capacity.
@@ -112,7 +94,12 @@ fvkmemory::resource_ptr<VulkanStage::Segment> VulkanStagePool::acquireStage(uint
uint32_t VulkanStagePool::alignToNonCoherentAtomSize(uint32_t bytes) {
VkDeviceSize alignment = mDeviceLimits->nonCoherentAtomSize;
return alignValue(bytes, alignment);
if (alignment == 0) {
return bytes;
}
uint32_t remainder = bytes % alignment;
return remainder == 0 ? bytes : bytes + (alignment - remainder);
}
VulkanStage* VulkanStagePool::allocateNewStage(uint32_t capacity) {
@@ -146,7 +133,7 @@ VulkanStage* VulkanStagePool::allocateNewStage(uint32_t capacity) {
#endif
}
return new VulkanStage(memory, buffer, capacity, pMapping);
return new VulkanStage(mAllocator, memory, buffer, capacity, pMapping);
}
void VulkanStagePool::destroyStage(VulkanStage const*&& stage) {
+16 -19
View File
@@ -35,8 +35,10 @@ class VulkanCommands;
// into smaller buffers as needed.
class VulkanStage {
public:
VulkanStage(VmaAllocation memory, VkBuffer buffer, uint32_t capacity, void* mapping)
: mMemory(memory),
VulkanStage(VmaAllocator allocator, VmaAllocation memory, VkBuffer buffer, uint32_t capacity,
void* mapping)
: mAllocator(allocator),
mMemory(memory),
mBuffer(buffer),
mCapacity(capacity),
mMapping(mapping) {}
@@ -51,10 +53,10 @@ public:
public:
using OnRecycle = std::function<void(uint32_t offset)>;
Segment(VulkanStage* parentStage, uint32_t capacity, uint32_t offset,
Segment(VmaAllocator allocator, VulkanStage* parentStage, uint32_t offset,
OnRecycle&& onRecycleFn)
: mParentStage(parentStage),
mCapacity(capacity),
: mAllocator(allocator),
mParentStage(parentStage),
mOffset(offset),
mOnRecycleFn(onRecycleFn) {}
@@ -70,12 +72,12 @@ public:
Segment& operator=(const Segment& other) = delete;
Segment& operator=(Segment&& other) = delete;
inline VulkanStage* parentStage() const { return mParentStage; }
inline VkBuffer buffer() const { return parentStage()->buffer(); }
inline VmaAllocation memory() const { return parentStage()->memory(); }
inline uint32_t capacity() const { return mCapacity; }
inline VkBuffer buffer() const { return mParentStage->buffer(); }
inline VmaAllocation memory() const { return mParentStage->memory(); }
inline uint32_t offset() const { return mOffset; }
VkResult copy(size_t dstOffset, void const* src, size_t writeSize);
inline void* mapping() const {
return reinterpret_cast<void*>(
reinterpret_cast<char*>(mParentStage->mapping()) + offset());
@@ -85,9 +87,9 @@ public:
// Ensure parent class can access the terminate method.
friend class VulkanStage;
VmaAllocator const mAllocator;
VulkanStage* const mParentStage;
const uint32_t mCapacity;
const uint32_t mOffset;
uint32_t const mOffset;
OnRecycle mOnRecycleFn;
};
@@ -105,13 +107,11 @@ public:
// Marks a region of the block as "in-use", and provides information about
// the allocated region to the caller. Note: this assumes that numBytes
// is aligned to the physical device's nonCoherentAtomSize.
// Takes the offset to use as the segment's start offset. It is assumed
// that this is greater than the segment's current offset, and it is
// undefined behavior if not.
fvkmemory::resource_ptr<Segment> acquireSegment(fvkmemory::ResourceManager* resManager,
uint32_t segmentOffset, uint32_t numBytes);
uint32_t numBytes);
private:
VmaAllocator const mAllocator;
const VmaAllocation mMemory;
const VkBuffer mBuffer;
const uint32_t mCapacity;
@@ -147,10 +147,7 @@ public:
// buffers so that we have less objects around that we have to keep track
// of.
// This function is NOT thread-safe.
// numBytes - the number of bytes required by this segment.
// alignment - the multiple to align the buffer offset to
fvkmemory::resource_ptr<VulkanStage::Segment> acquireStage(uint32_t numBytes,
uint32_t alignment = 0);
fvkmemory::resource_ptr<VulkanStage::Segment> acquireStage(uint32_t numBytes);
// Images have VK_IMAGE_LAYOUT_GENERAL and must not be transitioned to any other layout
VulkanStageImage const* acquireImage(PixelDataFormat format, PixelDataType type,
+16 -12
View File
@@ -219,8 +219,8 @@ VkImageUsageFlags getUsage(VulkanContext const& context, uint8_t samples,
return usage;
}
void adjustedMemcpy(void* mapped, PixelBufferDescriptor const& p, size_t width, size_t height,
size_t depth) {
void adjustedMemcpy(PixelBufferDescriptor const& p, size_t width, size_t height,
size_t depth, std::function<void(size_t, uint8_t*, size_t)> cpy) {
uint8_t* buf = (uint8_t*) p.buffer;
size_t const pixelSize = PixelBufferDescriptor::computeDataSize(p.format, p.type, 1, 1, 1);
size_t const pbdStride = p.stride ? p.stride : width;
@@ -243,13 +243,13 @@ void adjustedMemcpy(void* mapped, PixelBufferDescriptor const& p, size_t width,
for (size_t y = p.top; y < pbdHeight; y++) {
uint8_t* buf = (uint8_t*) p.buffer +
((p.left * pixelSize) + (y * pbdRowSize) + (z * pbdLayerSize));
uint8_t* curMapped = (uint8_t*) mapped + ((y - p.top) * rowSize + z * layerSize);
memcpy(curMapped, buf, writeSize);
uint32_t const offset = (y - p.top) * rowSize + z * layerSize;
cpy(offset, buf, writeSize);
}
}
} else {
size_t const writeSize = pixelSize * (width * height * depth);
memcpy(mapped, buf, writeSize);
cpy(0, buf, writeSize);
}
}
@@ -522,14 +522,15 @@ void VulkanTexture::updateImage(const PixelBufferDescriptor& data, uint32_t widt
// Note: the following stageSegment must be stored within the command buffer
// before going out of scope, to ensure proper bookkeeping within the
// staging buffer pool.
uint8_t alignment =
fvkutils::getTexelBlockSize(fvkutils::getVkFormat(hostData->format, hostData->type));
fvkmemory::resource_ptr<VulkanStage::Segment> stageSegment =
mState->mStagePool.acquireStage(writeSize, alignment);
mState->mStagePool.acquireStage(writeSize);
assert_invariant(stageSegment->memory());
adjustedMemcpy(stageSegment->mapping(), *hostData, width, height, depth);
vmaFlushAllocation(mState->mAllocator, stageSegment->memory(), stageSegment->offset(),
writeSize);
adjustedMemcpy(*hostData, width, height, depth,
[&stageSegment](size_t dstOffset, uint8_t* src, size_t numBytes) {
stageSegment->copy(dstOffset, src, numBytes);
});
// vmaFlushAllocation(mState->mAllocator, stageSegment->memory(), stageSegment->offset(),
// writeSize);
VulkanCommandBuffer& commands = mState->mCommands->get();
VkCommandBuffer const cmdbuf = commands.buffer();
@@ -599,7 +600,10 @@ void VulkanTexture::updateImageWithBlit(const PixelBufferDescriptor& data, uint3
VulkanStageImage const* stage
= mState->mStagePool.acquireImage(data.format, data.type, width, height);
vmaMapMemory(mState->mAllocator, stage->memory, &mapped);
adjustedMemcpy(mapped, data, width, height, depth);
adjustedMemcpy(data, width, height, depth,
[&mapped](size_t, uint8_t* src, size_t numBytes) {
memcpy(mapped, src, numBytes);
});
vmaUnmapMemory(mState->mAllocator, stage->memory);
vmaFlushAllocation(mState->mAllocator, stage->memory, 0, writeSize);
@@ -29,8 +29,6 @@ using namespace bluevk;
namespace filament::backend {
using namespace fvkutils;
VulkanYcbcrConversionCache::VulkanYcbcrConversionCache(VkDevice device)
: mDevice(device) {}
@@ -17,8 +17,6 @@
#ifndef TNT_FILAMENT_BACKEND_VULKANYCBCRCONVERSIONCACHE_H
#define TNT_FILAMENT_BACKEND_VULKANYCBCRCONVERSIONCACHE_H
#include "utils/Definitions.h"
#include <backend/DriverEnums.h>
#include <utils/Hash.h>
@@ -32,7 +30,7 @@ namespace filament::backend {
class VulkanYcbcrConversionCache {
public:
struct Params {
fvkutils::SamplerYcbcrConversion conversion = {}; // 4
SamplerYcbcrConversion conversion = {}; // 4
VkFormat format; // 4
uint64_t externalFormat = 0; // 8
};
@@ -47,15 +45,16 @@ private:
struct ConversionEqualTo {
bool operator()(Params lhs, Params rhs) const noexcept {
fvkutils::SamplerYcbcrConversion::EqualTo equal;
SamplerYcbcrConversion::EqualTo equal;
return equal(lhs.conversion, rhs.conversion) &&
lhs.externalFormat == rhs.externalFormat && lhs.format == rhs.format;
lhs.externalFormat == rhs.externalFormat &&
lhs.format == rhs.format;
}
};
using ConversionHashFn = utils::hash::MurmurHashFn<Params>;
tsl::robin_map<Params, VkSamplerYcbcrConversion, ConversionHashFn, ConversionEqualTo> mCache;
};
} // namespace filament::backend
}// namespace filament::backend
#endif// TNT_FILAMENT_BACKEND_VULKANYCBCRCONVERSIONCACHE_H
@@ -71,8 +71,6 @@ struct Resource {
restype(ResourceType::UNDEFINED_TYPE),
mHandleConsideredDestroyed(false) {}
uint32_t getCount() const { return mCount; }
private:
inline void inc() noexcept {
mCount++;
@@ -832,9 +832,6 @@ Driver* VulkanPlatform::createDriver(void* sharedContext,
context.mPhysicalDeviceFeatures.features.shaderClipDistance = VK_FALSE;
}
// TODO: Add support of VK_KHR_global_priority with `driverConfig.gpuContextPriority`
// in VulkanPlatform::createDriver.
ExtensionSet deviceExts;
// If using a shared context, we do not assume any extensions.
if (!mImpl->mSharedContext) {
@@ -926,7 +923,7 @@ Driver* VulkanPlatform::createDriver(void* sharedContext,
// Keep a copy of context for swapchains.
mImpl->mContext = context;
return VulkanDriver::create(this, mImpl->mContext, driverConfig);
return VulkanDriver::create(this, context, driverConfig);
}
// This needs to be explictly written for
@@ -233,212 +233,6 @@ VkFormat getVkFormat(TextureFormat format) {
}
}
// As per VUID-vkCmdCopyBufferToImage-dstImage-07975 and
// VUID-vkCmdCopyBufferToImage-dstImage-07978, these provide the texel block
// sizes for each format (as confirmed using the table at
// https://registry.khronos.org/vulkan/specs/latest/html/vkspec.html#formats-compatibility-classes).
// We have not listed values for multi-plane formats, as we do not support them
// unless they're externally provided.
uint8_t getTexelBlockSize(VkFormat format) {
switch (format) {
// 8-bit formats.
case VK_FORMAT_R8_UNORM:
case VK_FORMAT_R8_SNORM:
case VK_FORMAT_R8_USCALED:
case VK_FORMAT_R8_SSCALED:
case VK_FORMAT_R8_UINT:
case VK_FORMAT_R8_SINT:
case VK_FORMAT_R8_SRGB:
return 1;
// 16-bit formats.
case VK_FORMAT_R16_UNORM:
case VK_FORMAT_R16_SNORM:
case VK_FORMAT_R16_USCALED:
case VK_FORMAT_R16_SSCALED:
case VK_FORMAT_R16_UINT:
case VK_FORMAT_R16_SINT:
case VK_FORMAT_R16_SFLOAT:
case VK_FORMAT_R8G8_UNORM:
case VK_FORMAT_R8G8_SNORM:
case VK_FORMAT_R8G8_USCALED:
case VK_FORMAT_R8G8_SSCALED:
case VK_FORMAT_R8G8_UINT:
case VK_FORMAT_R8G8_SINT:
case VK_FORMAT_R8G8_SRGB:
case VK_FORMAT_R5G6B5_UNORM_PACK16:
case VK_FORMAT_R5G5B5A1_UNORM_PACK16:
case VK_FORMAT_R4G4B4A4_UNORM_PACK16:
return 2;
// 24-bit formats.
case VK_FORMAT_R8G8B8_UNORM:
case VK_FORMAT_R8G8B8_SNORM:
case VK_FORMAT_R8G8B8_USCALED:
case VK_FORMAT_R8G8B8_SSCALED:
case VK_FORMAT_R8G8B8_UINT:
case VK_FORMAT_R8G8B8_SINT:
case VK_FORMAT_R8G8B8_SRGB:
case VK_FORMAT_B8G8R8_UNORM:
case VK_FORMAT_B8G8R8_SNORM:
case VK_FORMAT_B8G8R8_USCALED:
case VK_FORMAT_B8G8R8_SSCALED:
case VK_FORMAT_B8G8R8_UINT:
case VK_FORMAT_B8G8R8_SINT:
case VK_FORMAT_B8G8R8_SRGB:
return 3;
// 32-bit formats.
case VK_FORMAT_R32_UINT:
case VK_FORMAT_R32_SINT:
case VK_FORMAT_R32_SFLOAT:
case VK_FORMAT_R16G16_UNORM:
case VK_FORMAT_R16G16_SNORM:
case VK_FORMAT_R16G16_USCALED:
case VK_FORMAT_R16G16_SSCALED:
case VK_FORMAT_R16G16_UINT:
case VK_FORMAT_R16G16_SINT:
case VK_FORMAT_R16G16_SFLOAT:
case VK_FORMAT_R8G8B8A8_UNORM:
case VK_FORMAT_R8G8B8A8_SNORM:
case VK_FORMAT_R8G8B8A8_USCALED:
case VK_FORMAT_R8G8B8A8_SSCALED:
case VK_FORMAT_R8G8B8A8_UINT:
case VK_FORMAT_R8G8B8A8_SINT:
case VK_FORMAT_R8G8B8A8_SRGB:
case VK_FORMAT_B8G8R8A8_UNORM:
case VK_FORMAT_B8G8R8A8_SNORM:
case VK_FORMAT_B8G8R8A8_USCALED:
case VK_FORMAT_B8G8R8A8_SSCALED:
case VK_FORMAT_B8G8R8A8_UINT:
case VK_FORMAT_B8G8R8A8_SINT:
case VK_FORMAT_B8G8R8A8_SRGB:
case VK_FORMAT_A8B8G8R8_UNORM_PACK32:
case VK_FORMAT_A8B8G8R8_SNORM_PACK32:
case VK_FORMAT_A8B8G8R8_USCALED_PACK32:
case VK_FORMAT_A8B8G8R8_SSCALED_PACK32:
case VK_FORMAT_A8B8G8R8_UINT_PACK32:
case VK_FORMAT_A8B8G8R8_SINT_PACK32:
case VK_FORMAT_A8B8G8R8_SRGB_PACK32:
case VK_FORMAT_A2B10G10R10_UNORM_PACK32:
// Depth and stencil formats.
case VK_FORMAT_S8_UINT:
case VK_FORMAT_D16_UNORM:
case VK_FORMAT_X8_D24_UNORM_PACK32:
case VK_FORMAT_D32_SFLOAT:
case VK_FORMAT_D24_UNORM_S8_UINT:
case VK_FORMAT_D32_SFLOAT_S8_UINT:
case VK_FORMAT_B10G11R11_UFLOAT_PACK32:
case VK_FORMAT_E5B9G9R9_UFLOAT_PACK32:
return 4;
// 48-bit formats.
case VK_FORMAT_R16G16B16_UNORM:
case VK_FORMAT_R16G16B16_SNORM:
case VK_FORMAT_R16G16B16_USCALED:
case VK_FORMAT_R16G16B16_SSCALED:
case VK_FORMAT_R16G16B16_UINT:
case VK_FORMAT_R16G16B16_SINT:
case VK_FORMAT_R16G16B16_SFLOAT:
return 6;
// 64-bit formats.
case VK_FORMAT_R32G32_UINT:
case VK_FORMAT_R32G32_SINT:
case VK_FORMAT_R32G32_SFLOAT:
case VK_FORMAT_R16G16B16A16_UNORM:
case VK_FORMAT_R16G16B16A16_SNORM:
case VK_FORMAT_R16G16B16A16_USCALED:
case VK_FORMAT_R16G16B16A16_SSCALED:
case VK_FORMAT_R16G16B16A16_UINT:
case VK_FORMAT_R16G16B16A16_SINT:
case VK_FORMAT_R16G16B16A16_SFLOAT:
// Compressed formats.
case VK_FORMAT_BC1_RGB_UNORM_BLOCK:
case VK_FORMAT_BC1_RGB_SRGB_BLOCK:
case VK_FORMAT_BC1_RGBA_UNORM_BLOCK:
case VK_FORMAT_BC1_RGBA_SRGB_BLOCK:
case VK_FORMAT_BC4_UNORM_BLOCK:
case VK_FORMAT_BC4_SNORM_BLOCK:
case VK_FORMAT_ETC2_R8G8B8_UNORM_BLOCK:
case VK_FORMAT_ETC2_R8G8B8_SRGB_BLOCK:
case VK_FORMAT_ETC2_R8G8B8A1_UNORM_BLOCK:
case VK_FORMAT_ETC2_R8G8B8A1_SRGB_BLOCK:
case VK_FORMAT_EAC_R11_UNORM_BLOCK:
case VK_FORMAT_EAC_R11_SNORM_BLOCK:
return 8;
// 96-bit formats.
case VK_FORMAT_R32G32B32_UINT:
case VK_FORMAT_R32G32B32_SINT:
case VK_FORMAT_R32G32B32_SFLOAT:
return 12;
// 128-bit formats.
case VK_FORMAT_R32G32B32A32_UINT:
case VK_FORMAT_R32G32B32A32_SINT:
case VK_FORMAT_R32G32B32A32_SFLOAT:
// Compressed formats.
case VK_FORMAT_BC2_UNORM_BLOCK:
case VK_FORMAT_BC2_SRGB_BLOCK:
case VK_FORMAT_BC3_UNORM_BLOCK:
case VK_FORMAT_BC3_SRGB_BLOCK:
case VK_FORMAT_BC5_UNORM_BLOCK:
case VK_FORMAT_BC5_SNORM_BLOCK:
case VK_FORMAT_BC6H_SFLOAT_BLOCK:
case VK_FORMAT_BC6H_UFLOAT_BLOCK:
case VK_FORMAT_BC7_UNORM_BLOCK:
case VK_FORMAT_BC7_SRGB_BLOCK:
case VK_FORMAT_ASTC_4x4_UNORM_BLOCK:
case VK_FORMAT_ASTC_5x4_UNORM_BLOCK:
case VK_FORMAT_ASTC_5x5_UNORM_BLOCK:
case VK_FORMAT_ASTC_6x5_UNORM_BLOCK:
case VK_FORMAT_ASTC_6x6_UNORM_BLOCK:
case VK_FORMAT_ASTC_8x5_UNORM_BLOCK:
case VK_FORMAT_ASTC_8x6_UNORM_BLOCK:
case VK_FORMAT_ASTC_8x8_UNORM_BLOCK:
case VK_FORMAT_ASTC_10x5_UNORM_BLOCK:
case VK_FORMAT_ASTC_10x6_UNORM_BLOCK:
case VK_FORMAT_ASTC_10x8_UNORM_BLOCK:
case VK_FORMAT_ASTC_10x10_UNORM_BLOCK:
case VK_FORMAT_ASTC_12x10_UNORM_BLOCK:
case VK_FORMAT_ASTC_12x12_UNORM_BLOCK:
case VK_FORMAT_ASTC_4x4_SRGB_BLOCK:
case VK_FORMAT_ASTC_5x4_SRGB_BLOCK:
case VK_FORMAT_ASTC_5x5_SRGB_BLOCK:
case VK_FORMAT_ASTC_6x5_SRGB_BLOCK:
case VK_FORMAT_ASTC_6x6_SRGB_BLOCK:
case VK_FORMAT_ASTC_8x5_SRGB_BLOCK:
case VK_FORMAT_ASTC_8x6_SRGB_BLOCK:
case VK_FORMAT_ASTC_8x8_SRGB_BLOCK:
case VK_FORMAT_ASTC_10x5_SRGB_BLOCK:
case VK_FORMAT_ASTC_10x6_SRGB_BLOCK:
case VK_FORMAT_ASTC_10x8_SRGB_BLOCK:
case VK_FORMAT_ASTC_10x10_SRGB_BLOCK:
case VK_FORMAT_ASTC_12x10_SRGB_BLOCK:
case VK_FORMAT_ASTC_12x12_SRGB_BLOCK:
case VK_FORMAT_ETC2_R8G8B8A8_UNORM_BLOCK:
case VK_FORMAT_ETC2_R8G8B8A8_SRGB_BLOCK:
case VK_FORMAT_EAC_R11G11_UNORM_BLOCK:
case VK_FORMAT_EAC_R11G11_SNORM_BLOCK:
return 16;
case VK_FORMAT_UNDEFINED:
// In cases where we've explicitly already determined that the
// format is not supported, let the rest of the system handle
// that explicitly. Return '1' for no alignment.
return 1;
default:
// This is the case where there's a format that we added support
// for in getVkFormat that we have not added support for in this
// function, and should be treated as an error.
assert_invariant(false && "Unknown data type, conversion is not supported.");
return 0;
}
}
VkFormat getVkFormat(PixelDataFormat format, PixelDataType type) {
if (type == PixelDataType::USHORT_565) return VK_FORMAT_R5G6B5_UNORM_PACK16;
if (type == PixelDataType::UINT_2_10_10_10_REV) return VK_FORMAT_A2B10G10R10_UNORM_PACK32;
@@ -17,8 +17,6 @@
#ifndef TNT_FILAMENT_BACKEND_VULKAN_UTILS_CONVERSION_H
#define TNT_FILAMENT_BACKEND_VULKAN_UTILS_CONVERSION_H
#include "Definitions.h"
#include <backend/DriverEnums.h>
#include <private/backend/BackendUtils.h> // for getFormatSize()
@@ -49,9 +47,6 @@ VkFormat getVkFormatLinear(VkFormat format);
// than a driver-level Texture format, and can account for a specified byte alignment.
uint32_t getBytesPerPixel(TextureFormat format);
// Required for aligning textures in staging buffers.
uint8_t getTexelBlockSize(VkFormat format);
VkCompareOp getCompareOp(SamplerCompareFunc func);
VkBlendFactor getBlendFactor(BlendFunction mode);
VkCullModeFlags getCullMode(CullingMode mode);
@@ -17,8 +17,6 @@
#ifndef TNT_FILAMENT_BACKEND_VULKAN_UTILS_DEFINITIONS_H
#define TNT_FILAMENT_BACKEND_VULKAN_UTILS_DEFINITIONS_H
#include <backend/DriverEnums.h>
#include <utils/bitset.h>
#include <utils/FixedCapacityVector.h>
@@ -381,81 +379,6 @@ static constexpr uint8_t getFragmentStageShift() noexcept {
// We have at most 4 descriptor sets. This is to indicate which ones are active.
using DescriptorSetMask = utils::bitset8;
//! this API is copied from (and only applies to) the Vulkan spec.
//! These specify YUV to RGB conversions.
enum class SamplerYcbcrModelConversion : uint8_t {
RGB_IDENTITY = 0,
YCBCR_IDENTITY = 1,
YCBCR_709 = 2,
YCBCR_601 = 3,
YCBCR_2020 = 4,
};
enum class SamplerYcbcrRange : uint8_t {
ITU_FULL = 0,
ITU_NARROW = 1,
};
enum class ChromaLocation : uint8_t {
COSITED_EVEN = 0,
MIDPOINT = 1,
};
//! Sampler parameters
struct SamplerYcbcrConversion { // NOLINT
SamplerYcbcrModelConversion ycbcrModel : 4;
TextureSwizzle r : 4;
TextureSwizzle g : 4;
TextureSwizzle b : 4;
TextureSwizzle a : 4;
SamplerYcbcrRange ycbcrRange : 1;
ChromaLocation xChromaOffset : 1;
ChromaLocation yChromaOffset : 1;
SamplerMagFilter chromaFilter : 1;
uint8_t padding;
struct Hasher {
size_t operator()(const SamplerYcbcrConversion p) const noexcept {
// we don't use std::hash<> here, so we don't have to include <functional>
return *reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&p));
}
};
struct EqualTo {
bool operator()(SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs) const noexcept {
assert_invariant(lhs.padding == 0);
auto* pLhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&lhs));
auto* pRhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&rhs));
return *pLhs == *pRhs;
}
};
struct LessThan {
bool operator()(SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs) const noexcept {
assert_invariant(lhs.padding == 0);
auto* pLhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&lhs));
auto* pRhs = reinterpret_cast<uint32_t const*>(reinterpret_cast<char const*>(&rhs));
return *pLhs < *pRhs;
}
};
private:
friend bool operator==(SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs) noexcept {
return SamplerYcbcrConversion::EqualTo{}(lhs, rhs);
}
friend bool operator!=(SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs) noexcept {
return !SamplerYcbcrConversion::EqualTo{}(lhs, rhs);
}
friend bool operator<(SamplerYcbcrConversion lhs, SamplerYcbcrConversion rhs) noexcept {
return SamplerYcbcrConversion::LessThan{}(lhs, rhs);
}
};
static_assert(sizeof(SamplerYcbcrConversion) == 4);
static_assert(sizeof(SamplerYcbcrConversion) <= sizeof(uint64_t),
"SamplerYcbcrConversion must be no more than 64 bits");
} // namespace filament::backend::fvkutils
#endif // TNT_FILAMENT_BACKEND_VULKAN_UTILS_DEFINITIONS_H
@@ -1,5 +0,0 @@
WebGPU does not include a trivial way to create mipmaps, and thus we must provide a render pass or compute pass to generate them.
This directory is the result of porting https://github.com/JolifantoBambla/webgpu-spd to C++, which itself is a WebGPU port of https://github.com/GPUOpen-LibrariesAndSDKs/FidelityFX-SDK/blob/main/docs/samples/single-pass-downsampler.md.
The first version of our port is primarily from a Gemini "Convert this to C++" request, which worked as an MVP with very few changes. Future commits will likely alter this version significantly, but it is being merged in this early state to facilitate rapid iteration.
@@ -1,956 +0,0 @@
#include "SpdMipmapGenerator.h"
#include <sstream>
#include <stdexcept>
// C++ port of https://github.com/JolifantoBambla/webgpu-spd for early experiments
namespace spd {
// Helper to convert enums to strings for map keys or shader code
const char* to_string(wgpu::TextureFormat format) {
switch (format) {
case wgpu::TextureFormat::RGBA8Unorm:
return "rgba8unorm";
case wgpu::TextureFormat::BGRA8Unorm:
return "bgra8unorm";
case wgpu::TextureFormat::R32Float:
return "r32float";
case wgpu::TextureFormat::RG32Float:
return "rg32float";
case wgpu::TextureFormat::RGBA32Float:
return "rgba32float";
case wgpu::TextureFormat::R16Float:
return "r16float";
case wgpu::TextureFormat::RG16Float:
return "rg16float";
case wgpu::TextureFormat::RGBA16Float:
return "rgba16float";
// Add other formats as needed
default:
return "rgba8unorm";
}
}
std::string MipmapGenerator::GetFilterCode(SPDFilter filter) {
switch (filter) {
case SPDFilter::Min:
return R"(
fn spd_reduce_4(v0: vec4<SPDScalar>, v1: vec4<SPDScalar>, v2: vec4<SPDScalar>, v3: vec4<SPDScalar>) -> vec4<SPDScalar> {
return min(min(v0, v1), min(v2, v3));
}
)";
case SPDFilter::Max:
return R"(
fn spd_reduce_4(v0: vec4<SPDScalar>, v1: vec4<SPDScalar>, v2: vec4<SPDScalar>, v3: vec4<SPDScalar>) -> vec4<SPDScalar> {
return max(max(v0, v1), max(v2, v3));
}
)";
case SPDFilter::MinMax:
return R"(
fn spd_reduce_4(v0: vec4<SPDScalar>, v1: vec4<SPDScalar>, v2: vec4<SPDScalar>, v3: vec4<SPDScalar>) -> vec4<SPDScalar> {
let max4 = max(max(v0.xy, v1.xy), max(v2.xy, v3.xy));
return vec4<SPDScalar>(min(min(v0.x, v1.x), min(v2.x, v3.x)), max(max4.x, max4.y), 0.0, 0.0);
}
)";
case SPDFilter::Average:
default:
return R"(
fn spd_reduce_4(v0: vec4<SPDScalar>, v1: vec4<SPDScalar>, v2: vec4<SPDScalar>, v3: vec4<SPDScalar>) -> vec4<SPDScalar> {
return (v0 + v1 + v2 + v3) * 0.25;
}
)";
}
}
MipmapGenerator::MipmapGenerator(const wgpu::Device& device)
: m_device(device) {
wgpu::BindGroupLayoutEntry bglEntry{};
bglEntry.binding = 0;
bglEntry.visibility = wgpu::ShaderStage::Compute;
bglEntry.buffer.type = wgpu::BufferBindingType::Uniform;
bglEntry.buffer.minBindingSize = 16;
wgpu::BindGroupLayoutDescriptor bglDesc{};
bglDesc.entryCount = 1;
bglDesc.entries = &bglEntry;
m_internalResourcesBindGroupLayout =
m_device.CreateBindGroupLayout(&bglDesc); // Basic BGL for <= 6 mips
wgpu::Limits deviceLimits;
m_device.GetLimits(&deviceLimits);
// Max number of *output* mips we can generate in a single pass.
// This is device.maxStorageTexturesPerShaderStage - 1 (for the source texture binding),
// further capped at 11 output mips (which means 12 total texture bindings: 1 source + 11
// outputs). WebGPU minspec for maxStorageTexturesPerShaderStage is 4, so
// deviceLimits.maxStorageTexturesPerShaderStage >= 4. Thus, m_maxMipsPerPass will be at least
// std::min(4u - 1, 11u) = 3.
m_maxMipsPerPass =
std::min(static_cast<uint32_t>(deviceLimits.maxStorageTexturesPerShaderStage) - 1, 11u);
m_maxArrayLayers = deviceLimits.maxStorageTexturesPerShaderStage;
// If we can generate more than 6 mips in a pass, we need an advanced BGL
// that includes atomic counter and mid-mip buffer for levels > 5.
if (m_maxMipsPerPass > 6) {
std::vector<wgpu::BindGroupLayoutEntry> entriesAdvanced(3);
// Entry 0: Uniform buffer (DownsamplePassMeta)
entriesAdvanced[0].binding = 0;
entriesAdvanced[0].visibility = wgpu::ShaderStage::Compute;
entriesAdvanced[0].buffer.type = wgpu::BufferBindingType::Uniform;
entriesAdvanced[0].buffer.minBindingSize = 16; // sizeof(DownsamplePassMeta) effectively
// Entry 1: Storage buffer (spd_global_counter)
entriesAdvanced[1].binding = 1;
entriesAdvanced[1].visibility = wgpu::ShaderStage::Compute;
entriesAdvanced[1].buffer.type = wgpu::BufferBindingType::Storage;
entriesAdvanced[1].buffer.minBindingSize = 4; // sizeof(atomic<u32>)
// Entry 2: Storage buffer (mip_dst_6_buffer)
entriesAdvanced[2].binding = 2;
entriesAdvanced[2].visibility = wgpu::ShaderStage::Compute;
entriesAdvanced[2].buffer.type = wgpu::BufferBindingType::Storage;
// Size for array<array<array<vec4<f32>, 64>, 64>> which is 16 (bytes for vec4f32) * 64 * 64
// * arrayLayers. The shader defines it as [slice][uv.y][uv.x], so 64x64 per slice.
// minBindingSize here is for one slice as a baseline. Actual buffer will be larger for
// multiple array layers.
entriesAdvanced[2].buffer.minBindingSize = 16 * 64 * 64;
wgpu::BindGroupLayoutDescriptor bglDescAdvanced{};
bglDescAdvanced.label = "SPD Internal BGL (Advanced)";
bglDescAdvanced.entryCount = static_cast<uint32_t>(entriesAdvanced.size());
bglDescAdvanced.entries = entriesAdvanced.data();
m_internalResourcesBindGroupLayout_Advanced =
m_device.CreateBindGroupLayout(&bglDescAdvanced);
}
}
void MipmapGenerator::PreparePipelines(wgpu::TextureFormat format, SPDFilter filter,
bool halfPrecision) {
SPDScalarType scalarType = SanitizeScalarType(format, halfPrecision);
// Prepare for a reasonable number of mips, up to m_maxMipsPerPass
// The loop should go from 1 up to and including m_maxMipsPerPass.
// If m_maxMipsPerPass is 0 (e.g. maxStorageTexturesPerShaderStage <=1 ), this loop won't run.
for (uint32_t i = 1; i <= m_maxMipsPerPass; ++i) {
GetOrCreatePipeline({ format, scalarType, filter, i });
}
}
SPDScalarType MipmapGenerator::SanitizeScalarType(wgpu::TextureFormat format, bool halfPrecision) {
std::string formatStr = to_string(format);
std::transform(formatStr.begin(), formatStr.end(), formatStr.begin(), ::tolower);
SPDScalarType texelType = SPDScalarType::F32; // Default to F32
if (formatStr.find("sint") != std::string::npos) {
texelType = SPDScalarType::I32;
} else if (formatStr.find("uint") != std::string::npos) {
texelType = SPDScalarType::U32;
}
if (halfPrecision && texelType == SPDScalarType::F32) {
if (m_device.HasFeature(wgpu::FeatureName::ShaderF16)) {
return SPDScalarType::F16;
} else {
// Optional: Log a warning here that halfPrecision was requested for a float format
// but the device does not support ShaderF16, so F32 will be used.
}
} else if (halfPrecision && texelType != SPDScalarType::F32) {
// Optional: Log a warning here that halfPrecision was requested for a non-float format.
}
return texelType;
}
SPDPipeline& MipmapGenerator::GetOrCreatePipeline(const PipelineCacheKey& key) {
uint32_t numMipsForPipeline = std::min(key.numMips, m_maxMipsPerPass);
if (m_pipelines.find(key) == m_pipelines.end()) {
// Create the pipeline
SPDPipeline spdPipeline;
std::vector<wgpu::BindGroupLayoutEntry> mipsBglEntries;
// Bindings are 0 for source, 1 to numMipsForPipeline for destinations
for (uint32_t i = 0; i <= numMipsForPipeline; ++i) {
wgpu::BindGroupLayoutEntry entry{};
entry.binding = i;
entry.visibility = wgpu::ShaderStage::Compute;
if (i == 0) {
entry.texture.sampleType = (key.scalarType == SPDScalarType::I32)
? wgpu::TextureSampleType::Sint
: (key.scalarType == SPDScalarType::U32)
? wgpu::TextureSampleType::Uint
: wgpu::TextureSampleType::UnfilterableFloat;
entry.texture.viewDimension = wgpu::TextureViewDimension::e2DArray;
} else {
entry.storageTexture.access = wgpu::StorageTextureAccess::WriteOnly;
entry.storageTexture.format = key.format;
entry.storageTexture.viewDimension = wgpu::TextureViewDimension::e2DArray;
}
mipsBglEntries.push_back(entry);
}
wgpu::BindGroupLayoutDescriptor mipsBglDesc{};
mipsBglDesc.entryCount = mipsBglEntries.size();
mipsBglDesc.entries = mipsBglEntries.data();
spdPipeline.mipsBindGroupLayout = m_device.CreateBindGroupLayout(&mipsBglDesc);
std::string shaderCode = MakeShaderCode(key.format, GetFilterCode(key.filter),
numMipsForPipeline, key.scalarType);
wgpu::ShaderModuleWGSLDescriptor wgslDesc{};
wgslDesc.code = shaderCode.c_str();
wgpu::ShaderModuleDescriptor shaderModuleDesc{};
shaderModuleDesc.nextInChain = &wgslDesc;
wgpu::ShaderModule shaderModule = m_device.CreateShaderModule(&shaderModuleDesc);
wgpu::BindGroupLayout chosenInternalBGL =
(numMipsForPipeline > 6 && m_internalResourcesBindGroupLayout_Advanced)
? m_internalResourcesBindGroupLayout_Advanced
: m_internalResourcesBindGroupLayout;
wgpu::BindGroupLayout bgls[] = { spdPipeline.mipsBindGroupLayout, chosenInternalBGL };
wgpu::PipelineLayoutDescriptor layoutDesc{};
layoutDesc.bindGroupLayoutCount = 2;
layoutDesc.bindGroupLayouts = bgls;
wgpu::ComputePipelineDescriptor pipelineDesc{};
pipelineDesc.layout = m_device.CreatePipelineLayout(&layoutDesc);
pipelineDesc.compute.module = shaderModule;
pipelineDesc.compute.entryPoint = "downsample";
spdPipeline.pipeline = m_device.CreateComputePipeline(&pipelineDesc);
m_pipelines[key] = std::move(spdPipeline);
}
return m_pipelines[key];
}
void MipmapGenerator::Generate(wgpu::CommandEncoder& commandEncoder, wgpu::Texture srcTexture,
const SPDPassConfig& config) {
if (config.numMips <= 1 || config.numMips <= config.sourceMipLevel + 1) {
return;
}
if (m_maxMipsPerPass == 0) {
return;
}
PreparePipelines(srcTexture.GetFormat(), config.filter, config.halfPrecision);
uint32_t totalSrcArrayLayers = srcTexture.GetDepthOrArrayLayers();
// Loop over array layer chunks
for (uint32_t currentBaseArrayLayer = 0; currentBaseArrayLayer < totalSrcArrayLayers;
currentBaseArrayLayer += m_maxArrayLayers) {
uint32_t arrayLayersThisPass =
std::min(m_maxArrayLayers, totalSrcArrayLayers - currentBaseArrayLayer);
uint32_t mipsToGenerateCount = config.numMips - (config.sourceMipLevel + 1);
uint32_t currentSourceMipLevel = config.sourceMipLevel;
while (mipsToGenerateCount > 0) {
uint32_t mipsThisPass = std::min(mipsToGenerateCount, m_maxMipsPerPass);
SPDPassConfig passSpecificConfig = {};
passSpecificConfig.filter = config.filter;
passSpecificConfig.halfPrecision = config.halfPrecision;
passSpecificConfig.targetTexture =
config.targetTexture ? config.targetTexture : srcTexture;
passSpecificConfig.numMips = mipsThisPass;
passSpecificConfig.sourceMipLevel = currentSourceMipLevel;
// Pass baseArrayLayer and numArrayLayers directly to generatePass
generatePass(commandEncoder, srcTexture, passSpecificConfig, currentBaseArrayLayer,
arrayLayersThisPass);
mipsToGenerateCount -= mipsThisPass;
currentSourceMipLevel += mipsThisPass;
}
}
}
void MipmapGenerator::generatePass(wgpu::CommandEncoder& commandEncoder, wgpu::Texture srcTexture,
const SPDPassConfig& passConfig, uint32_t baseArrayLayer, uint32_t numArrayLayers) {
uint32_t width = srcTexture.GetWidth() >> passConfig.sourceMipLevel;
uint32_t height = srcTexture.GetHeight() >> passConfig.sourceMipLevel;
uint32_t arrayLayerCountForThisPass = numArrayLayers;
wgpu::Texture target = passConfig.targetTexture ? passConfig.targetTexture : srcTexture;
uint32_t numMipsThisPass = passConfig.numMips;
if (numMipsThisPass == 0) return;
SPDScalarType scalarType = SanitizeScalarType(srcTexture.GetFormat(), passConfig.halfPrecision);
SPDPipeline& spdPipeline = GetOrCreatePipeline(
{ target.GetFormat(), scalarType, passConfig.filter, numMipsThisPass });
// --- Create Bind Group 0 (Mips) ---
std::vector<wgpu::BindGroupEntry> mipEntries;
wgpu::TextureViewDescriptor srcViewDesc{};
srcViewDesc.dimension = wgpu::TextureViewDimension::e2DArray;
srcViewDesc.baseMipLevel = passConfig.sourceMipLevel;
srcViewDesc.mipLevelCount = 1;
srcViewDesc.baseArrayLayer = baseArrayLayer;
srcViewDesc.arrayLayerCount = arrayLayerCountForThisPass;
wgpu::BindGroupEntry srcEntry{};
srcEntry.binding = 0;
srcEntry.textureView = srcTexture.CreateView(&srcViewDesc);
mipEntries.push_back(srcEntry);
for (uint32_t i = 0; i < numMipsThisPass; ++i) {
wgpu::TextureViewDescriptor dstViewDesc{};
dstViewDesc.dimension = wgpu::TextureViewDimension::e2DArray;
dstViewDesc.baseMipLevel = passConfig.sourceMipLevel + i + 1;
dstViewDesc.mipLevelCount = 1;
dstViewDesc.baseArrayLayer = baseArrayLayer;
dstViewDesc.arrayLayerCount = arrayLayerCountForThisPass;
wgpu::BindGroupEntry dstEntry{};
dstEntry.binding = i + 1;
dstEntry.textureView = target.CreateView(&dstViewDesc);
mipEntries.push_back(dstEntry);
}
wgpu::BindGroupDescriptor mipBindGroupDesc{};
mipBindGroupDesc.layout = spdPipeline.mipsBindGroupLayout;
mipBindGroupDesc.entryCount = mipEntries.size();
mipBindGroupDesc.entries = mipEntries.data();
wgpu::BindGroup mipsBindGroup = m_device.CreateBindGroup(&mipBindGroupDesc);
// --- Create Bind Group 1 (Internal Resources) ---
uint32_t widthDiv64 = (width + 63) / 64;
uint32_t heightDiv64 = (height + 63) / 64;
uint32_t numWorkGroups = widthDiv64 * heightDiv64;
struct DownsamplePassMeta {
uint32_t work_group_offset[2] = { 0, 0 };
uint32_t num_work_groups;
uint32_t mips;
uint32_t padding[12];
} meta;
meta.num_work_groups = numWorkGroups;
meta.mips = numMipsThisPass;
wgpu::BufferDescriptor metaBufferDesc{};
metaBufferDesc.size = sizeof(DownsamplePassMeta);
metaBufferDesc.usage = wgpu::BufferUsage::Uniform | wgpu::BufferUsage::CopyDst;
wgpu::Buffer metaBuffer = m_device.CreateBuffer(&metaBufferDesc);
m_device.GetQueue().WriteBuffer(metaBuffer, 0, &meta, sizeof(meta));
wgpu::BindGroup internalBindGroup;
wgpu::BindGroupLayout chosenInternalBGL =
(numMipsThisPass > 6 && m_internalResourcesBindGroupLayout_Advanced)
? m_internalResourcesBindGroupLayout_Advanced
: m_internalResourcesBindGroupLayout;
if (numMipsThisPass > 6 && m_internalResourcesBindGroupLayout_Advanced) {
wgpu::BufferDescriptor atomicCounterBufferDesc{};
atomicCounterBufferDesc.size = sizeof(uint32_t) * arrayLayerCountForThisPass;
atomicCounterBufferDesc.usage = wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopyDst;
wgpu::Buffer atomicCounterBuffer = m_device.CreateBuffer(&atomicCounterBufferDesc);
commandEncoder.ClearBuffer(atomicCounterBuffer, 0, WGPU_WHOLE_SIZE);
wgpu::BufferDescriptor midMipBufferDesc{};
midMipBufferDesc.size = 64 * 64 * 16 * arrayLayerCountForThisPass;
midMipBufferDesc.usage = wgpu::BufferUsage::Storage;
wgpu::Buffer midMipBuffer = m_device.CreateBuffer(&midMipBufferDesc);
std::vector<wgpu::BindGroupEntry> internalEntries(3);
internalEntries[0].binding = 0;
internalEntries[0].buffer = metaBuffer;
internalEntries[0].size = sizeof(DownsamplePassMeta);
internalEntries[1].binding = 1;
internalEntries[1].buffer = atomicCounterBuffer;
internalEntries[1].size = atomicCounterBuffer.GetSize();
internalEntries[2].binding = 2;
internalEntries[2].buffer = midMipBuffer;
internalEntries[2].size = midMipBuffer.GetSize();
wgpu::BindGroupDescriptor internalBindGroupDesc{};
internalBindGroupDesc.layout = chosenInternalBGL;
internalBindGroupDesc.entryCount = static_cast<uint32_t>(internalEntries.size());
internalBindGroupDesc.entries = internalEntries.data();
internalBindGroup = m_device.CreateBindGroup(&internalBindGroupDesc);
} else {
wgpu::BindGroupEntry metaEntry{};
metaEntry.binding = 0;
metaEntry.buffer = metaBuffer;
metaEntry.size = sizeof(DownsamplePassMeta);
wgpu::BindGroupDescriptor internalBindGroupDesc{};
internalBindGroupDesc.layout = chosenInternalBGL;
internalBindGroupDesc.entryCount = 1;
internalBindGroupDesc.entries = &metaEntry;
internalBindGroup = m_device.CreateBindGroup(&internalBindGroupDesc);
}
// --- Dispatch ---
wgpu::ComputePassEncoder pass = commandEncoder.BeginComputePass();
pass.SetPipeline(spdPipeline.pipeline);
pass.SetBindGroup(0, mipsBindGroup);
pass.SetBindGroup(1, internalBindGroup);
pass.DispatchWorkgroups(widthDiv64, heightDiv64, arrayLayerCountForThisPass);
pass.End();
}
// Main shader generation logic
// Helper function to check if a string is in a vector of strings
bool includes(const std::vector<std::string>& vec, const std::string& str) {
for (const auto& s: vec) {
if (s == str) {
return true;
}
}
return false;
}
std::string MakeShaderCode(wgpu::TextureFormat outputFormat, const std::string& filterOp,
unsigned int numMips, // Assuming a default value for numMips
SPDScalarType scalarType) { // Default scalarType
std::stringstream ss;
std::string texelType;
if (scalarType == SPDScalarType::I32) {
texelType = "i32";
} else if (scalarType == SPDScalarType::U32) {
texelType = "u32";
} else {
texelType = "f32";
}
bool useF16 = (scalarType == SPDScalarType::F16);
std::string filterCode = filterOp;
if (filterOp == SPD_FILTER_AVERAGE && !includes({ "f32", "f16" }, texelType)) {
// Replace "* 0.25" with "/ 4"
size_t pos = filterCode.find("* 0.25");
if (pos != std::string::npos) {
filterCode.replace(pos, std::string("* 0.25").length(), "/ 4");
}
}
// Generate mipsBindings
std::string mipsBindings;
for (unsigned int i = 0; i < numMips; ++i) {
mipsBindings += "@group(0) @binding(" + std::to_string(i + 1) + ") var dst_mip_" +
std::to_string(i + 1) + ": texture_storage_2d_array<" +
to_string(outputFormat) + ", write>;\n";
}
// Generate mipsAccessorBody
std::string mipsAccessorBody;
for (unsigned int i = 0; i < numMips; ++i) {
if (i == 5 && numMips > 6) {
mipsAccessorBody += " else if mip == 6 {\n";
mipsAccessorBody += " textureStore(dst_mip_6, uv, slice, " +
(useF16 ? "vec4<" + texelType + ">(value)" : "value") + ");\n";
mipsAccessorBody += " mip_dst_6_buffer[slice][uv.y][uv.x] = value;\n";
mipsAccessorBody += " }";
} else {
if (i != 0) {
mipsAccessorBody += " else ";
}
mipsAccessorBody += "if mip == " + std::to_string(i + 1) + " {\n";
mipsAccessorBody += " textureStore(dst_mip_" + std::to_string(i + 1) +
", uv, slice, " +
(useF16 ? "vec4<" + texelType + ">(value)" : "value") + ");\n";
mipsAccessorBody += " }";
}
}
std::string mipsAccessor =
"fn store_dst_mip(value: vec4<SPDScalar>, uv: vec2<u32>, slice: u32, mip: u32) {\n" +
mipsAccessorBody + "\n}";
std::string midMipAccessor = "return mip_dst_6_buffer[slice][uv.y][uv.x];";
// Start building the final shader code string
ss << R"(
// This file is part of the FidelityFX SDK.
//
// Copyright (C) 2023 Advanced Micro Devices, Inc.
//
// Permission is hereby granted, free of charge, to any person obtaining a copy
// of this software and associated documentation files(the “Software”), to deal
// in the Software without restriction, including without limitation the rights
// to use, copy, modify, merge, publish, distribute, sublicense, and /or sell
// copies of the Software, and to permit persons to whom the Software is
// furnished to do so, subject to the following conditions :
//
// The above copyright notice and this permission notice shall be included in
// all copies or substantial portions of the Software.
//
// THE SOFTWARE IS PROVIDED “AS IS”, WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.IN NO EVENT SHALL THE
// AUTHORS COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
// THE SOFTWARE.
// Definitions --------------------------------------------------------------------------------------------------------
)";
if (useF16) {
ss << "enable f16;\n";
}
ss << "alias SPDScalar = " << texelType
<< ";\n\n"; // Using texelType here, assuming SPDScalar maps to it.
ss << R"(
// Helpers ------------------------------------------------------------------------------------------------------------
/**
* A helper function performing a remap 64x1 to 8x8 remapping which is necessary for 2D wave reductions.
* * The 64-wide lane indices to 8x8 remapping is performed as follows:
* 00 01 08 09 10 11 18 19
* 02 03 0a 0b 12 13 1a 1b
* 04 05 0c 0d 14 15 1c 1d
* 06 07 0e 0f 16 17 1e 1f
* 20 21 28 29 30 31 38 39
* 22 23 2a 2b 32 33 3a 3b
* 24 25 2c 2d 34 35 3c 3d
* 26 27 2e 2f 36 37 3e 3f
* * @param a: The input 1D coordinate to remap.
*
* @returns The remapped 2D coordinates.
*/
fn remap_for_wave_reduction(a: u32) -> vec2<u32> {
return vec2<u32>(
insertBits(extractBits(a, 2u, 3u), a, 0u, 1u),
insertBits(extractBits(a, 3u, 3u), extractBits(a, 1u, 2u), 0u, 2u)
);
}
fn map_to_xy(local_invocation_index: u32) -> vec2<u32> {
let sub_xy: vec2<u32> = remap_for_wave_reduction(local_invocation_index % 64);
return vec2<u32>(
sub_xy.x + 8 * ((local_invocation_index >> 6) % 2),
sub_xy.y + 8 * ((local_invocation_index >> 7))
);
}
/*
* Compute a linear value from a SRGB value.
* * @param value: The value to convert to linear from SRGB.
* * @returns A value in SRGB space.
*/
/*
fn srgb_to_linear(value: SPDScalar) -> SPDScalar {
let j = vec3<SPDScalar>(0.0031308 * 12.92, 12.92, 1.0 / 2.4);
let k = vec2<SPDScalar>(1.055, -0.055);
return clamp(j.x, value * j.y, pow(value, j.z) * k.x + k.y);
}
*/
// Resources & Accessors -----------------------------------------------------------------------------------------------
struct DownsamplePassMeta {
work_group_offset: vec2<u32>,
num_work_groups: u32,
mips: u32,
}
// In the original version dst_mip_i is an image2Darray [SPD_MAX_MIP_LEVELS+1], i.e., 12+1, but WGSL doesn't support arrays of textures yet
// Also these are read_write because for mips 7-13, the workgroup reads from mip level 6 - since most formats don't support read_write access in WGSL yet, we use a single read_write buffer in such cases instead
@group(0) @binding(0) var src_mip_0: texture_2d_array<)"
<< texelType << R"(>;
)" << mipsBindings
<< R"(
@group(1) @binding(0) var<uniform> downsample_pass_meta : DownsamplePassMeta;
@group(1) @binding(1) var<storage, read_write> spd_global_counter: array<atomic<u32>>;
@group(1) @binding(2) var<storage, read_write> mip_dst_6_buffer: array<array<array<vec4<f32>, 64>, 64>>;
fn get_mips() -> u32 {
return downsample_pass_meta.mips;
}
fn get_num_work_groups() -> u32 {
return downsample_pass_meta.num_work_groups;
}
fn get_work_group_offset() -> vec2<u32> {
return downsample_pass_meta.work_group_offset;
}
fn load_src_image(uv: vec2<u32>, slice: u32) -> vec4<SPDScalar> {
return vec4<SPDScalar>(textureLoad(src_mip_0, uv, slice, 0));
}
fn load_mid_mip_image(uv: vec2<u32>, slice: u32) -> vec4<SPDScalar> {
)";
if (numMips > 6) {
ss << midMipAccessor;
} else {
ss << "return vec4<SPDScalar>();";
}
ss << R"(
}
)" << mipsAccessor
<< R"(
// Workgroup -----------------------------------------------------------------------------------------------------------
var<workgroup> spd_intermediate: array<array<vec4<SPDScalar>, 16>, 16>;
var<workgroup> spd_counter: atomic<u32>;
fn spd_increase_atomic_counter(slice: u32) {
atomicStore(&spd_counter, atomicAdd(&spd_global_counter[slice], 1));
}
fn spd_get_atomic_counter() -> u32 {
return atomicLoad(&spd_counter);
}
fn spd_reset_atomic_counter(slice: u32) {
atomicStore(&spd_global_counter[slice], 0);
}
// Cotnrol flow --------------------------------------------------------------------------------------------------------
fn spd_barrier() {
// in glsl this does: groupMemoryBarrier(); barrier();
workgroupBarrier();
}
// Only last active workgroup should proceed
fn spd_exit_workgroup(num_work_groups: u32, local_invocation_index: u32, slice: u32) -> bool {
// global atomic counter
if (local_invocation_index == 0) {
spd_increase_atomic_counter(slice);
}
spd_barrier();
return spd_get_atomic_counter() != (num_work_groups - 1);
}
// Pixel access --------------------------------------------------------------------------------------------------------
)" << filterCode
<< R"(
fn spd_store(pix: vec2<u32>, out_value: vec4<SPDScalar>, mip: u32, slice: u32) {
store_dst_mip(out_value, pix, slice, mip + 1);
}
fn spd_load_intermediate(x: u32, y: u32) -> vec4<SPDScalar> {
return spd_intermediate[x][y];
}
fn spd_store_intermediate(x: u32, y: u32, value: vec4<SPDScalar>) {
spd_intermediate[x][y] = value;
}
fn spd_reduce_intermediate(i0: vec2<u32>, i1: vec2<u32>, i2: vec2<u32>, i3: vec2<u32>) -> vec4<SPDScalar> {
let v0 = spd_load_intermediate(i0.x, i0.y);
let v1 = spd_load_intermediate(i1.x, i1.y);
let v2 = spd_load_intermediate(i2.x, i2.y);
let v3 = spd_load_intermediate(i3.x, i3.y);
return spd_reduce_4(v0, v1, v2, v3);
}
fn spd_reduce_load_4(base: vec2<u32>, slice: u32) -> vec4<SPDScalar> {
let v0 = load_src_image(base + vec2<u32>(0, 0), slice);
let v1 = load_src_image(base + vec2<u32>(0, 1), slice);
let v2 = load_src_image(base + vec2<u32>(1, 0), slice);
let v3 = load_src_image(base + vec2<u32>(1, 1), slice);
return spd_reduce_4(v0, v1, v2, v3);
}
fn spd_reduce_load_mid_mip_4(base: vec2<u32>, slice: u32) -> vec4<SPDScalar> {
let v0 = load_mid_mip_image(base + vec2<u32>(0, 0), slice);
let v1 = load_mid_mip_image(base + vec2<u32>(0, 1), slice);
let v2 = load_mid_mip_image(base + vec2<u32>(1, 0), slice);
let v3 = load_mid_mip_image(base + vec2<u32>(1, 1), slice);
return spd_reduce_4(v0, v1, v2, v3);
}
// Main logic ---------------------------------------------------------------------------------------------------------
fn spd_downsample_mips_0_1(x: u32, y: u32, workgroup_id: vec2<u32>, local_invocation_index: u32, mip: u32, slice: u32) {
var v: array<vec4<SPDScalar>, 4>;
let workgroup64 = workgroup_id.xy * 64;
let workgroup32 = workgroup_id.xy * 32;
let workgroup16 = workgroup_id.xy * 16;
var tex = workgroup64 + vec2<u32>(x * 2, y * 2);
var pix = workgroup32 + vec2<u32>(x, y);
v[0] = spd_reduce_load_4(tex, slice);
spd_store(pix, v[0], 0, slice);
tex = workgroup64 + vec2<u32>(x * 2 + 32, y * 2);
pix = workgroup32 + vec2<u32>(x + 16, y);
v[1] = spd_reduce_load_4(tex, slice);
spd_store(pix, v[1], 0, slice);
tex = workgroup64 + vec2<u32>(x * 2, y * 2 + 32);
pix = workgroup32 + vec2<u32>(x, y + 16);
v[2] = spd_reduce_load_4(tex, slice);
spd_store(pix, v[2], 0, slice);
tex = workgroup64 + vec2<u32>(x * 2 + 32, y * 2 + 32);
pix = workgroup32 + vec2<u32>(x + 16, y + 16);
v[3] = spd_reduce_load_4(tex, slice);
spd_store(pix, v[3], 0, slice);
if mip <= 1 {
return;
}
for (var i = 0u; i < 4u; i++) {
spd_store_intermediate(x, y, v[i]);
spd_barrier();
if local_invocation_index < 64 {
v[i] = spd_reduce_intermediate(
vec2<u32>(x * 2 + 0, y * 2 + 0),
vec2<u32>(x * 2 + 1, y * 2 + 0),
vec2<u32>(x * 2 + 0, y * 2 + 1),
vec2<u32>(x * 2 + 1, y * 2 + 1)
);
spd_store(workgroup16 + vec2<u32>(x + (i % 2) * 8, y + (i / 2) * 8), v[i], 1, slice);
}
spd_barrier();
}
if local_invocation_index < 64 {
spd_store_intermediate(x + 0, y + 0, v[0]);
spd_store_intermediate(x + 8, y + 0, v[1]);
spd_store_intermediate(x + 0, y + 8, v[2]);
spd_store_intermediate(x + 8, y + 8, v[3]);
}
}
fn spd_downsample_mip_2(x: u32, y: u32, workgroup_id: vec2<u32>, local_invocation_index: u32, mip: u32, slice: u32) {
if local_invocation_index < 64u {
let v = spd_reduce_intermediate(
vec2<u32>(x * 2 + 0, y * 2 + 0),
vec2<u32>(x * 2 + 1, y * 2 + 0),
vec2<u32>(x * 2 + 0, y * 2 + 1),
vec2<u32>(x * 2 + 1, y * 2 + 1)
);
spd_store(workgroup_id.xy * 8 + vec2<u32>(x, y), v, mip, slice);
// store to LDS, try to reduce bank conflicts
// x 0 x 0 x 0 x 0 x 0 x 0 x 0 x 0
// 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0
// 0 x 0 x 0 x 0 x 0 x 0 x 0 x 0 x
// 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0
// x 0 x 0 x 0 x 0 x 0 x 0 x 0 x 0
// ...
// x 0 x 0 x 0 x 0 x 0 x 0 x 0 x 0
spd_store_intermediate(x * 2 + y % 2, y * 2, v);
}
}
fn spd_downsample_mip_3(x: u32, y: u32, workgroup_id: vec2<u32>, local_invocation_index: u32, mip: u32, slice: u32) {
if local_invocation_index < 16u {
// x 0 x 0
// 0 0 0 0
// 0 x 0 x
// 0 0 0 0
let v = spd_reduce_intermediate(
vec2<u32>(x * 4 + 0 + 0, y * 4 + 0),
vec2<u32>(x * 4 + 2 + 0, y * 4 + 0),
vec2<u32>(x * 4 + 0 + 1, y * 4 + 2),
vec2<u32>(x * 4 + 2 + 1, y * 4 + 2)
);
spd_store(workgroup_id.xy * 4 + vec2<u32>(x, y), v, mip, slice);
// store to LDS
// x 0 0 0 x 0 0 0 x 0 0 0 x 0 0 0
// 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0
// 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0
// 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0 0
// 0 x 0 0 0 x 0 0 0 x 0 0 0 x 0 0
// ...
// 0 0 x 0 0 0 x 0 0 0 x 0 0 0 x 0
// ...
// 0 0 0 x 0 0 0 x 0 0 0 x 0 0 0 x
// ...
spd_store_intermediate(x * 4 + y, y * 4, v);
}
}
fn spd_downsample_mip_4(x: u32, y: u32, workgroup_id: vec2<u32>, local_invocation_index: u32, mip: u32, slice: u32) {
if local_invocation_index < 4u {
// x 0 0 0 x 0 0 0
// ...
// 0 x 0 0 0 x 0 0
let v = spd_reduce_intermediate(
vec2<u32>(x * 8 + 0 + 0 + y * 2, y * 8 + 0),
vec2<u32>(x * 8 + 4 + 0 + y * 2, y * 8 + 0),
vec2<u32>(x * 8 + 0 + 1 + y * 2, y * 8 + 4),
vec2<u32>(x * 8 + 4 + 1 + y * 2, y * 8 + 4)
);
spd_store(workgroup_id.xy * 2 + vec2<u32>(x, y), v, mip, slice);
// store to LDS
// x x x x 0 ...
// 0 ...
spd_store_intermediate(x + y * 2, 0, v);
}
}
fn spd_downsample_mip_5(workgroup_id: vec2<u32>, local_invocation_index: u32, mip: u32, slice: u32) {
if local_invocation_index < 1u {
// x x x x 0 ...
// 0 ...
let v = spd_reduce_intermediate(vec2<u32>(0, 0), vec2<u32>(1, 0), vec2<u32>(2, 0), vec2<u32>(3, 0));
spd_store(workgroup_id.xy, v, mip, slice);
}
}
fn spd_downsample_next_four(x: u32, y: u32, workgroup_id: vec2<u32>, local_invocation_index: u32, base_mip: u32, mips: u32, slice: u32) {
if mips <= base_mip {
return;
}
spd_barrier();
spd_downsample_mip_2(x, y, workgroup_id, local_invocation_index, base_mip, slice);
if mips <= base_mip + 1 {
return;
}
spd_barrier();
spd_downsample_mip_3(x, y, workgroup_id, local_invocation_index, base_mip + 1, slice);
if mips <= base_mip + 2 {
return;
}
spd_barrier();
spd_downsample_mip_4(x, y, workgroup_id, local_invocation_index, base_mip + 2, slice);
if mips <= base_mip + 3 {
return;
}
spd_barrier();
spd_downsample_mip_5(workgroup_id, local_invocation_index, base_mip + 3, slice);
}
fn spd_downsample_last_four(x: u32, y: u32, workgroup_id: vec2<u32>, local_invocation_index: u32, base_mip: u32, mips: u32, slice: u32, exit: bool) {
if mips <= base_mip {
return;
}
spd_barrier();
if !exit {
spd_downsample_mip_2(x, y, workgroup_id, local_invocation_index, base_mip, slice);
}
if mips <= base_mip + 1 {
return;
}
spd_barrier();
if !exit {
spd_downsample_mip_3(x, y, workgroup_id, local_invocation_index, base_mip + 1, slice);
}
if mips <= base_mip + 2 {
return;
}
spd_barrier();
if !exit {
spd_downsample_mip_4(x, y, workgroup_id, local_invocation_index, base_mip + 2, slice);
}
if mips <= base_mip + 3 {
return;
}
spd_barrier();
if !exit {
spd_downsample_mip_5(workgroup_id, local_invocation_index, base_mip + 3, slice);
}
}
fn spd_downsample_mips_6_7(x: u32, y: u32, mips: u32, slice: u32) {
var tex = vec2<u32>(x * 4 + 0, y * 4 + 0);
var pix = vec2<u32>(x * 2 + 0, y * 2 + 0);
let v0 = spd_reduce_load_mid_mip_4(tex, slice);
spd_store(pix, v0, 6, slice);
tex = vec2<u32>(x * 4 + 2, y * 4 + 0);
pix = vec2<u32>(x * 2 + 1, y * 2 + 0);
let v1 = spd_reduce_load_mid_mip_4(tex, slice);
spd_store(pix, v1, 6, slice);
tex = vec2<u32>(x * 4 + 0, y * 4 + 2);
pix = vec2<u32>(x * 2 + 0, y * 2 + 1);
let v2 = spd_reduce_load_mid_mip_4(tex, slice);
spd_store(pix, v2, 6, slice);
tex = vec2<u32>(x * 4 + 2, y * 4 + 2);
pix = vec2<u32>(x * 2 + 1, y * 2 + 1);
let v3 = spd_reduce_load_mid_mip_4(tex, slice);
spd_store(pix, v3, 6, slice);
if mips <= 7 {
return;
}
// no barrier needed, working on values only from the same thread
let v = spd_reduce_4(v0, v1, v2, v3);
spd_store(vec2<u32>(x, y), v, 7, slice);
spd_store_intermediate(x, y, v);
}
fn spd_downsample_last_6(x: u32, y: u32, local_invocation_index: u32, mips: u32, num_work_groups: u32, slice: u32) {
if mips <= 6 {
return;
}
// increase the global atomic counter for the given slice and check if it's the last remaining thread group:
// terminate if not, continue if yes.
let exit = spd_exit_workgroup(num_work_groups, local_invocation_index, slice);
// can't exit directly because subsequent barrier calls break uniform control flow...
if !exit {
// reset the global atomic counter back to 0 for the next spd dispatch
spd_reset_atomic_counter(slice);
// After mip 5 there is only a single workgroup left that downsamples the remaining up to 64x64 texels.
// compute MIP level 6 and 7
spd_downsample_mips_6_7(x, y, mips, slice);
}
// compute MIP level 8, 9, 10, 11
spd_downsample_last_four(x, y, vec2<u32>(0, 0), local_invocation_index, 8, mips, slice, exit);
}
/// Downsamples a 64x64 tile based on the work group id.
/// If after downsampling it's the last active thread group, computes the remaining MIP levels.
///
/// @param [in] workGroupID index of the work group / thread group
/// @param [in] localInvocationIndex index of the thread within the thread group in 1D
/// @param [in] mips the number of total MIP levels to compute for the input texture
/// @param [in] numWorkGroups the total number of dispatched work groups / thread groups for this slice
/// @param [in] slice the slice of the input texture
fn spd_downsample(workgroup_id: vec2<u32>, local_invocation_index: u32, mips: u32, num_work_groups: u32, slice: u32) {
let xy = map_to_xy(local_invocation_index);
spd_downsample_mips_0_1(xy.x, xy.y, workgroup_id, local_invocation_index, mips, slice);
spd_downsample_next_four(xy.x, xy.y, workgroup_id, local_invocation_index, 2, mips, slice);
)";
if (numMips > 6) {
ss << " spd_downsample_last_6(xy.x, xy.y, local_invocation_index, mips, "
"num_work_groups, slice);\n";
}
ss << R"(}
// Entry points -------------------------------------------------------------------------------------------------------
@compute
@workgroup_size(256, 1, 1)
fn downsample(@builtin(local_invocation_index) local_invocation_index: u32, @builtin(workgroup_id) workgroup_id: vec3<u32>) {
spd_downsample(
workgroup_id.xy + get_work_group_offset(),
local_invocation_index,
get_mips(),
get_num_work_groups(),
workgroup_id.z
);
}
)";
return ss.str();
}
} // namespace spd
@@ -1,99 +0,0 @@
#pragma once
#include <optional>
#include <string>
#include <tsl/robin_map.h>
#include <vector>
#include <webgpu/webgpu_cpp.h>
// C++ port of https://github.com/JolifantoBambla/webgpu-spd for early experiments
namespace spd {
// Enum for selecting the downsampling filter.
enum class SPDFilter { Average, Min, Max, MinMax };
// Enum for shader scalar types.
enum class SPDScalarType { F32, F16, I32, U32 };
// Configuration for a single mipmap generation pass.
struct SPDPassConfig {
SPDFilter filter = SPDFilter::Average;
wgpu::Texture targetTexture = nullptr;
uint32_t numMips = 0; // For the public API, this is total desired mips. Internally, mips for
// the current pass.
bool halfPrecision = false;
uint32_t sourceMipLevel = 0;
};
// Holds a pipeline and its corresponding bind group layout.
struct SPDPipeline {
wgpu::BindGroupLayout mipsBindGroupLayout = nullptr;
wgpu::ComputePipeline pipeline = nullptr;
};
// Manages pipeline creation, caching, and execution for mipmap generation.
class MipmapGenerator {
public:
MipmapGenerator(const wgpu::Device& device);
// Pre-creates pipelines for specified formats and filters.
void PreparePipelines(wgpu::TextureFormat format, SPDFilter filter, bool halfPrecision = false);
// Generates a compute pass for creating mipmaps.
void Generate(wgpu::CommandEncoder& commandEncoder, wgpu::Texture srcTexture,
const SPDPassConfig& config);
private:
void generatePass(wgpu::CommandEncoder& commandEncoder, wgpu::Texture srcTexture,
const SPDPassConfig& passConfig, uint32_t baseArrayLayer,
uint32_t numArrayLayers); // Internal pass config might differ slightly if needed
wgpu::Device m_device;
wgpu::BindGroupLayout m_internalResourcesBindGroupLayout;
wgpu::BindGroupLayout m_internalResourcesBindGroupLayout_Advanced = nullptr;
// Maximum number of mips that can be generated in a single pass.
uint32_t m_maxMipsPerPass;
uint32_t m_maxArrayLayers;
// Key for the pipeline cache.
struct PipelineCacheKey {
wgpu::TextureFormat format;
SPDScalarType scalarType;
SPDFilter filter;
uint32_t numMips;
bool operator==(const PipelineCacheKey& other) const {
return format == other.format && scalarType == other.scalarType &&
filter == other.filter && numMips == other.numMips;
}
};
// Hash function for PipelineCacheKey.
struct PipelineCacheKeyHash {
std::size_t operator()(const PipelineCacheKey& key) const {
std::size_t h1 = std::hash<wgpu::TextureFormat>()(key.format);
std::size_t h2 = std::hash<SPDScalarType>()(key.scalarType);
std::size_t h3 = std::hash<SPDFilter>()(key.filter);
std::size_t h4 = std::hash<uint32_t>()(key.numMips);
// A simple way to combine hashes.
return h1 ^ (h2 << 1) ^ (h3 << 2) ^ (h4 << 3);
}
};
// Cached pipelines.
tsl::robin_map<PipelineCacheKey, SPDPipeline, PipelineCacheKeyHash> m_pipelines;
// Helper methods
SPDPipeline& GetOrCreatePipeline(const PipelineCacheKey& key);
SPDScalarType SanitizeScalarType(wgpu::TextureFormat format, bool halfPrecision);
std::string GetFilterCode(SPDFilter filter);
};
// Assuming SPD_FILTER_AVERAGE is a string constant
const std::string SPD_FILTER_AVERAGE = "value * 0.25"; // Original filter operation
// Generates the WGSL shader code dynamically.
std::string MakeShaderCode(wgpu::TextureFormat outputFormat,
const std::string& filterOp = SPD_FILTER_AVERAGE, uint32_t numMips = 0,
SPDScalarType scalarType = SPDScalarType::F32);
} // namespace spd
@@ -1,21 +0,0 @@
MIT License
Copyright (c) 2024 Lukas Herzberger
Permission is hereby granted, free of charge, to any person obtaining a copy
of this software and associated documentation files (the "Software"), to deal
in the Software without restriction, including without limitation the rights
to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
copies of the Software, and to permit persons to whom the Software is
furnished to do so, subject to the following conditions:
The above copyright notice and this permission notice shall be included in all
copies or substantial portions of the Software.
THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
SOFTWARE.
@@ -117,10 +117,10 @@ WebGPUDescriptorSetLayout::WebGPUDescriptorSetLayout(DescriptorSetLayout const&
samplerEntryInfo.binding = samplerEntry.binding;
samplerEntry.visibility = wEntry.visibility;
wEntry.texture.multisampled = isMultiSampledTypeDescriptor(fEntry.type);
// TODO: Set once we have the filtering values
if (isDepthDescriptor(fEntry.type)) {
samplerEntry.sampler.type = wgpu::SamplerBindingType::Comparison;
} else if (isIntDescriptor(fEntry.type) ||
any(fEntry.flags & DescriptorFlags::UNFILTERABLE)) {
} else if (isIntDescriptor(fEntry.type)) {
samplerEntry.sampler.type = wgpu::SamplerBindingType::NonFiltering;
} else {
samplerEntry.sampler.type = wgpu::SamplerBindingType::Filtering;
@@ -151,11 +151,8 @@ WebGPUDescriptorSetLayout::WebGPUDescriptorSetLayout(DescriptorSetLayout const&
if (isDepthDescriptor(fEntry.type)) {
wEntry.texture.sampleType = wgpu::TextureSampleType::Depth;
} else if (isFloatDescriptor(fEntry.type)) {
if (any(fEntry.flags & DescriptorFlags::UNFILTERABLE)) {
wEntry.texture.sampleType = wgpu::TextureSampleType::UnfilterableFloat;
} else {
wEntry.texture.sampleType = wgpu::TextureSampleType::Float;
}
// TODO: Set once we have the filtering values
wEntry.texture.sampleType = wgpu::TextureSampleType::Float;
} else if (isIntDescriptor(fEntry.type)) {
wEntry.texture.sampleType = wgpu::TextureSampleType::Sint;
} else if (isUnsignedIntDescriptor(fEntry.type)) {
+132 -201
View File
@@ -18,7 +18,6 @@
#include "WebGPUBufferObject.h"
#include "WebGPUDescriptorSet.h"
#include "WebGPUDescriptorSetLayout.h"
#include "WebGPUFence.h"
#include "WebGPUIndexBuffer.h"
#include "WebGPUPipelineCreation.h"
#include "WebGPUProgram.h"
@@ -54,8 +53,6 @@
#include <sstream>
#include <utility>
using namespace std::chrono_literals;
namespace filament::backend {
Driver* WebGPUDriver::create(WebGPUPlatform& platform, const Platform::DriverConfig& driverConfig) noexcept {
@@ -67,14 +64,13 @@ Driver* WebGPUDriver::create(WebGPUPlatform& platform, const Platform::DriverCon
WebGPUDriver::WebGPUDriver(WebGPUPlatform& platform,
const Platform::DriverConfig& driverConfig) noexcept
: mPlatform{ platform },
mAdapter{ mPlatform.requestAdapter(nullptr) },
mDevice{ mPlatform.requestDevice(mAdapter) },
mQueue{ mDevice.GetQueue() },
mMipMapGenerator{ mDevice },
mHandleAllocator{ "Handles", driverConfig.handleArenaSize,
driverConfig.disableHandleUseAfterFreeCheck, driverConfig.disableHeapHandleTags } {
: mPlatform(platform),
mHandleAllocator("Handles", driverConfig.handleArenaSize,
driverConfig.disableHandleUseAfterFreeCheck, driverConfig.disableHeapHandleTags) {
mAdapter = mPlatform.requestAdapter(nullptr);
mDevice = mPlatform.requestDevice(mAdapter);
mDevice.GetLimits(&mDeviceLimits);
mQueue = mDevice.GetQueue();
}
WebGPUDriver::~WebGPUDriver() noexcept = default;
@@ -104,7 +100,6 @@ void WebGPUDriver::terminate() {
void WebGPUDriver::tick(int) {
mDevice.Tick();
mAdapter.GetInstance().ProcessEvents();
}
void WebGPUDriver::beginFrame(int64_t monotonic_clock_ns,
@@ -266,9 +261,7 @@ Handle<HwProgram> WebGPUDriver::createProgramS() noexcept {
}
Handle<HwFence> WebGPUDriver::createFenceS() noexcept {
// The handle must be constructed here, as a synchronous call to get the status
// might happen before createFenceR is executed.
return allocAndConstructHandle<WebGPUFence, HwFence>();
return Handle<HwFence>((Handle<HwFence>::HandleId) mNextFakeHandle++);
}
Handle<HwTimerQuery> WebGPUDriver::createTimerQueryS() noexcept {
@@ -341,6 +334,9 @@ void WebGPUDriver::createSwapChainR(Handle<HwSwapChain> sch, void* nativeWindow,
mDevice, flags);
assert_invariant(mSwapChain);
FWGPU_LOGW << "WebGPU support is highly experimental, in development, and tested for only a "
"small set of simple samples (e.g. hellotriangle and texturedquad), thus issues "
"are likely to be encountered at this stage.";
#if !FWGPU_ENABLED(FWGPU_PRINT_SYSTEM) && !defined(NDEBUG)
char printSystemHex[16];
snprintf(printSystemHex, sizeof(printSystemHex), "%#x", FWGPU_PRINT_SYSTEM);
@@ -452,18 +448,15 @@ void WebGPUDriver::createDefaultRenderTargetR(Handle<HwRenderTarget> renderTarge
}
void WebGPUDriver::createRenderTargetR(Handle<HwRenderTarget> renderTargetHandle,
const TargetBufferFlags targetFlags, const uint32_t width, const uint32_t height,
const TargetBufferFlags targets, const uint32_t width, const uint32_t height,
const uint8_t samples, const uint8_t layerCount, const MRT color,
const TargetBufferInfo depth, const TargetBufferInfo stencil) {
constructHandle<WebGPURenderTarget>(renderTargetHandle, width, height, samples, layerCount,
color, depth, stencil, targetFlags);
color, depth, stencil);
}
void WebGPUDriver::createFenceR(Handle<HwFence> fenceHandle, const int /* dummy */) {
// the handle was already constructed in createFenceS
const auto fence = handleCast<WebGPUFence>(fenceHandle);
assert_invariant(mQueue);
fence->addMarkerToQueueState(mQueue);
void WebGPUDriver::createFenceR(Handle<HwFence> fh, int) {
//todo
}
void WebGPUDriver::createTimerQueryR(Handle<HwTimerQuery> tqh, int) {}
@@ -516,18 +509,13 @@ void WebGPUDriver::updateStreams(CommandStream* driver) {
//todo
}
void WebGPUDriver::destroyFence(Handle<HwFence> fenceHandle) {
if (fenceHandle) {
destructHandle<WebGPUFence>(fenceHandle);
}
void WebGPUDriver::destroyFence(Handle<HwFence> fh) {
//todo
}
FenceStatus WebGPUDriver::getFenceStatus(Handle<HwFence> fenceHandle) {
const auto fence = handleCast<WebGPUFence>(fenceHandle);
if (!fence) {
return FenceStatus::ERROR;
}
return fence->getStatus();
FenceStatus WebGPUDriver::getFenceStatus(Handle<HwFence> fh) {
//todo
return FenceStatus::CONDITION_SATISFIED;
}
// We create all textures using VK_IMAGE_TILING_OPTIMAL, so our definition of "supported" is that
@@ -617,28 +605,15 @@ uint8_t WebGPUDriver::getMaxDrawBuffers() {
}
size_t WebGPUDriver::getMaxUniformBufferSize() {
return mDeviceLimits.maxUniformBufferBindingSize;
return 16384u;
}
size_t WebGPUDriver::getMaxTextureSize(const SamplerType target) {
size_t result = 2048u;
switch (target) {
case SamplerType::SAMPLER_2D:
case SamplerType::SAMPLER_2D_ARRAY:
case SamplerType::SAMPLER_EXTERNAL:
case SamplerType::SAMPLER_CUBEMAP:
case SamplerType::SAMPLER_CUBEMAP_ARRAY:
result = mDeviceLimits.maxTextureDimension2D;
break;
case SamplerType::SAMPLER_3D:
result = mDeviceLimits.maxTextureDimension3D;
break;
}
return result;
return 2048u;
}
size_t WebGPUDriver::getMaxArrayTextureLayers() {
return mDeviceLimits.maxTextureArrayLayers;
return 256u;
}
void WebGPUDriver::updateIndexBuffer(Handle<HwIndexBuffer> indexBufferHandle,
@@ -763,46 +738,67 @@ void WebGPUDriver::setExternalStream(Handle<HwTexture> textureHandle,
}
void WebGPUDriver::generateMipmaps(Handle<HwTexture> textureHandle) {
if (!mCommandEncoder) {
mMipQueue.push_back(textureHandle);
return;
}
auto texture = handleCast<WebGPUTexture>(textureHandle);
assert_invariant(texture);
wgpu::Texture wgpuTexture = texture->getTexture();
assert_invariant(wgpuTexture);
const auto usage = wgpuTexture.GetUsage();
FILAMENT_CHECK_PRECONDITION(usage & wgpu::TextureUsage::TextureBinding)
<< "Texture for mipmap generation must have TextureBinding usage. "
"supportsMultipleMipLevels for texture format "
<< static_cast<uint32_t>(wgpuTexture.GetFormat()) << " is "
<< WebGPUTexture::supportsMultipleMipLevelsViaStorageBinding(wgpuTexture.GetFormat());
FILAMENT_CHECK_PRECONDITION(usage & wgpu::TextureUsage::StorageBinding)
<< "Texture for mipmap generation must have StorageBinding usage. "
"supportsMultipleMipLevels for texture format "
<< static_cast<uint32_t>(wgpuTexture.GetFormat()) << " is "
<< WebGPUTexture::supportsMultipleMipLevelsViaStorageBinding(wgpuTexture.GetFormat());
FILAMENT_CHECK_PRECONDITION(wgpuTexture.GetUsage() & wgpu::TextureUsage::CopySrc)
<< "Texture intended for mipmap generation (as source) must have CopySrc usage.";
FILAMENT_CHECK_PRECONDITION(wgpuTexture.GetUsage() & wgpu::TextureUsage::CopyDst)
<< "Texture intended for mipmap generation (as destination) must have CopyDst usage.";
const uint32_t totalMipLevels = wgpuTexture.GetMipLevelCount();
if (totalMipLevels <= 1) {
uint32_t mipLevelCount = wgpuTexture.GetMipLevelCount();
if (mipLevelCount <= 1) {
return;
}
FILAMENT_CHECK_PRECONDITION(texture->supportsMultipleMipLevels())
<< "Calling generateMipmaps(...) on a texture that doesn't support generating them. "
"sampler type "
<< to_string(texture->target) << " levels " << static_cast<uint32_t>(texture->levels)
<< " wgpu format (enum value) "
<< static_cast<uint32_t>(texture->getTexture().GetFormat()) << " width "
<< texture->width << " height " << texture->height << " depth " << texture->depth
<< " usage flags (as a number) " << static_cast<uint32_t>(texture->usage);
// We will record all passes into a single command encoder.
wgpu::CommandEncoderDescriptor encoderDesc = {};
encoderDesc.label = "Mipmap Command Encoder";
wgpu::CommandEncoder encoder = mDevice.CreateCommandEncoder(&encoderDesc);
uint32_t width = wgpuTexture.GetWidth();
uint32_t height = wgpuTexture.GetHeight();
// For 3D textures, depth is > 1. For 2D/Cube/Array, effectively 1 for mip-level copies.
uint32_t depth =
(texture->target == SamplerType::SAMPLER_3D) ? wgpuTexture.GetDepthOrArrayLayers() : 1;
spd::SPDPassConfig spdConfig = { .filter = spd::SPDFilter::Average,
.targetTexture = wgpuTexture,
.numMips = totalMipLevels,
.halfPrecision = false,
.sourceMipLevel = 0 };
for (uint32_t mipLevel = 0; mipLevel < mipLevelCount - 1; ++mipLevel) {
wgpu::TexelCopyTextureInfo sourceCopyInfo{
.texture = wgpuTexture,
.mipLevel = mipLevel,
.aspect = texture->getAspect(),
};
mMipMapGenerator.Generate(encoder, wgpuTexture, spdConfig);
wgpu::TexelCopyTextureInfo destinationCopyInfo{
.texture = wgpuTexture,
.mipLevel = mipLevel + 1,
.aspect = texture->getAspect(),
};
// Finish the encoder and submit all the passes at once.
wgpu::CommandBufferDescriptor cmdBufferDesc = {};
cmdBufferDesc.label = "Mipmap Command Buffer";
wgpu::CommandBuffer commandBuffer = encoder.Finish(&cmdBufferDesc);
mQueue.Submit(1, &commandBuffer);
uint32_t dstWidth = std::max(1u, width >> 1);
uint32_t dstHeight = std::max(1u, height >> 1);
uint32_t dstDepth = std::max(1u, depth >> 1);
wgpu::Extent3D copySize{ .width = dstWidth,
.height = dstHeight,
.depthOrArrayLayers = (texture->target == SamplerType::SAMPLER_3D)
? dstDepth
: texture->getArrayLayerCount() };
mCommandEncoder.CopyTextureToTexture(&sourceCopyInfo, &destinationCopyInfo, &copySize);
width = dstWidth;
height = dstHeight;
depth = dstDepth;
}
}
void WebGPUDriver::compilePrograms(CompilerPriorityQueue priority,
@@ -823,9 +819,10 @@ void WebGPUDriver::beginRenderPass(Handle<HwRenderTarget> renderTargetHandle,
std::array<wgpu::TextureView, MRT::MAX_SUPPORTED_RENDER_TARGET_COUNT> customColorViews{};
uint32_t customColorViewCount = 0;
wgpu::TextureView customDepthStencilView = nullptr;
wgpu::TextureFormat customDepthStencilFormat = wgpu::TextureFormat::Undefined;
wgpu::TextureView customDepthView = nullptr;
wgpu::TextureFormat customDepthFormat = wgpu::TextureFormat::Undefined;
wgpu::TextureFormat customStencilFormat = wgpu::TextureFormat::Undefined;
wgpu::TextureView customStencilView = nullptr;
mCurrentRenderTarget = renderTarget;
if (renderTarget->isDefaultRenderTarget()) {
@@ -833,24 +830,6 @@ void WebGPUDriver::beginRenderPass(Handle<HwRenderTarget> renderTargetHandle,
defaultColorView = mTextureView;
defaultDepthStencilView = mSwapChain->getDepthTextureView();
defaultDepthStencilFormat = mSwapChain->getDepthFormat();
if (any(renderTarget->getTargetFlags() & TargetBufferFlags::STENCIL) &&
!(defaultDepthStencilFormat == wgpu::TextureFormat::Depth24PlusStencil8 ||
defaultDepthStencilFormat == wgpu::TextureFormat::Depth32FloatStencil8 ||
defaultDepthStencilFormat == wgpu::TextureFormat::Stencil8)) {
FILAMENT_CHECK_POSTCONDITION(false) << "Default render target requested stencil, but swap chain's depth format "
<< (uint32_t)defaultDepthStencilFormat << " does not have a stencil aspect.";
}
if (any(renderTarget->getTargetFlags() & TargetBufferFlags::DEPTH) &&
!(defaultDepthStencilFormat == wgpu::TextureFormat::Depth16Unorm ||
defaultDepthStencilFormat == wgpu::TextureFormat::Depth32Float ||
defaultDepthStencilFormat == wgpu::TextureFormat::Depth24Plus ||
defaultDepthStencilFormat == wgpu::TextureFormat::Depth24PlusStencil8 ||
defaultDepthStencilFormat == wgpu::TextureFormat::Depth32FloatStencil8)) {
FILAMENT_CHECK_POSTCONDITION(false) << "Default render target requested depth, but swap chain's depth format "
<< (uint32_t)defaultDepthStencilFormat << " does not have a depth aspect.";
}
} else {
// Resolve views for custom render target
const auto& colorInfos = renderTarget->getColorAttachmentInfos();
@@ -870,63 +849,36 @@ void WebGPUDriver::beginRenderPass(Handle<HwRenderTarget> renderTargetHandle,
}
}
}
const auto& depthInfo = renderTarget->getDepthAttachmentInfo();
const auto& stencilInfo = renderTarget->getStencilAttachmentInfo();
Handle<HwTexture> depthStencilSourceHandle = {};
uint8_t depthStencilMipLevel = 0;
uint32_t depthStencilArrayLayer = 0;
if (depthInfo.handle && stencilInfo.handle) {
auto const depthTexture{handleCast<WebGPUTexture>(depthInfo.handle)->getTexture()};
auto const stencilTexture{handleCast<WebGPUTexture>(stencilInfo.handle)->getTexture()};
FILAMENT_CHECK_POSTCONDITION(depthTexture.Get() == stencilTexture.Get())
<< "Filament must reference the same resource if both depth and stencil "
"textures are present";
}
if (depthInfo.handle) {
depthStencilSourceHandle = depthInfo.handle;
depthStencilMipLevel = depthInfo.level;
depthStencilArrayLayer = depthInfo.layer;
} else if (stencilInfo.handle) {
depthStencilSourceHandle = stencilInfo.handle;
depthStencilMipLevel = stencilInfo.level;
depthStencilArrayLayer = stencilInfo.layer;
FILAMENT_CHECK_POSTCONDITION(depthInfo.layer < renderTarget->getLayerCount())
<< "Depth attachment requests layer " << depthInfo.layer
<< "but render target has only " << renderTarget->getLayerCount() << ".";
auto depthTexture = handleCast<WebGPUTexture>(depthInfo.handle);
if (depthTexture) {
const uint8_t depthMipLevel = depthInfo.level;
const uint32_t depthArrayLayer = depthInfo.layer;
customDepthView =
depthTexture->getOrMakeTextureView(depthMipLevel, depthArrayLayer);
customDepthFormat = depthTexture->getViewFormat();
}
}
if (depthStencilSourceHandle) {
auto dsTexture = handleCast<WebGPUTexture>(depthStencilSourceHandle);
if (dsTexture) {
customDepthStencilView = dsTexture->getOrMakeTextureView(depthStencilMipLevel, depthStencilArrayLayer);
customDepthStencilFormat = dsTexture->getViewFormat();
if (any(renderTarget->getTargetFlags() & TargetBufferFlags::STENCIL) &&
!(customDepthStencilFormat == wgpu::TextureFormat::Depth24PlusStencil8 ||
customDepthStencilFormat ==
wgpu::TextureFormat::Depth32FloatStencil8 ||
customDepthStencilFormat ==
wgpu::TextureFormat::Stencil8)) {
FILAMENT_CHECK_POSTCONDITION(false)
<< "Custom render target requested stencil, but the provided texture"
"format number"
<< (uint32_t) customDepthStencilFormat
<< " does not have a stencil aspect.";
}
if (any(renderTarget->getTargetFlags() & TargetBufferFlags::DEPTH) &&
!(customDepthStencilFormat == wgpu::TextureFormat::Depth16Unorm ||
customDepthStencilFormat == wgpu::TextureFormat::Depth32Float ||
customDepthStencilFormat ==
wgpu::TextureFormat::Depth24Plus ||
customDepthStencilFormat ==
wgpu::TextureFormat::Depth24PlusStencil8 ||
customDepthStencilFormat ==
wgpu::TextureFormat::Depth32FloatStencil8)) {
FILAMENT_CHECK_POSTCONDITION(false) << "Custom render target requested depth, "
"but the provided texture format number"
<< (uint32_t) customDepthStencilFormat
<< " does not have a depth aspect.";
}
const auto& stencilInfo = renderTarget->getStencilAttachmentInfo();
if (stencilInfo.handle) {
// If depth and stencil use the same texture handle, this will re-cast but that's fine.
auto stencilTexture = handleCast<WebGPUTexture>(stencilInfo.handle);
if (stencilTexture) {
FILAMENT_CHECK_POSTCONDITION(stencilInfo.layer < renderTarget->getLayerCount())
<< "Stencil attachment requests layer " << stencilInfo.layer
<< " but render target has only " << renderTarget->getLayerCount()
<< " layers.";
const uint8_t stencilMipLevel = stencilInfo.level;
const uint32_t stencilArrayLayer = stencilInfo.layer;
customStencilView =
stencilTexture->getOrMakeTextureView(stencilMipLevel, stencilArrayLayer);
customStencilFormat = stencilTexture->getViewFormat();
}
}
}
@@ -935,9 +887,13 @@ void WebGPUDriver::beginRenderPass(Handle<HwRenderTarget> renderTargetHandle,
params,
defaultColorView,
defaultDepthStencilView,
defaultDepthStencilFormat,
customColorViews.data(),
customColorViewCount,
customDepthStencilView);
customDepthView,
customStencilView,
customDepthFormat,
customStencilFormat);
mRenderPassEncoder = mCommandEncoder.BeginRenderPass(&renderPassDescriptor);
@@ -972,26 +928,16 @@ void WebGPUDriver::makeCurrent(Handle<HwSwapChain> drawSch, Handle<HwSwapChain>
wgpu::Extent2D surfaceSize = mPlatform.getSurfaceExtent(mNativeWindow);
mTextureView = mSwapChain->getCurrentSurfaceTextureView(surfaceSize);
assert_invariant(mTextureView);
assert_invariant(mDefaultRenderTarget);
wgpu::TextureFormat depthFormat = mSwapChain->getDepthFormat();
TargetBufferFlags newTargetFlags = filament::backend::TargetBufferFlags::NONE;
//Assuming Color and Depth are always present.
newTargetFlags |= filament::backend::TargetBufferFlags::COLOR;
if (depthFormat != wgpu::TextureFormat::Undefined) {
newTargetFlags |= filament::backend::TargetBufferFlags::DEPTH;
if (depthFormat == wgpu::TextureFormat::Depth24PlusStencil8 ||
depthFormat == wgpu::TextureFormat::Depth32FloatStencil8) {
newTargetFlags |= filament::backend::TargetBufferFlags::STENCIL;
}
}
mDefaultRenderTarget->setTargetFlags(newTargetFlags);
wgpu::CommandEncoderDescriptor commandEncoderDescriptor = { .label = "frame_command_encoder" };
wgpu::CommandEncoderDescriptor commandEncoderDescriptor = {
.label = "command_encoder"
};
mCommandEncoder = mDevice.CreateCommandEncoder(&commandEncoderDescriptor);
if (!mMipQueue.empty()) {
for (auto& handle: mMipQueue) {
generateMipmaps(handle);
}
mMipQueue.clear();
}
assert_invariant(mCommandEncoder);
}
@@ -1003,22 +949,6 @@ void WebGPUDriver::commit(Handle<HwSwapChain> sch) {
assert_invariant(mCommandBuffer);
mCommandEncoder = nullptr;
mQueue.Submit(1, &mCommandBuffer);
static bool firstRender = true;
// For the first frame rendered, we need to make sure the work is done before presenting or we
// get a purple flash
if (firstRender) {
auto f = mQueue.OnSubmittedWorkDone(wgpu::CallbackMode::WaitAnyOnly,
[=](wgpu::QueueWorkDoneStatus) {});
const wgpu::Instance instance = mAdapter.GetInstance();
auto wStatus = instance.WaitAny(f,
std::chrono::duration_cast<std::chrono::nanoseconds>(1s).count());
if (wStatus != wgpu::WaitStatus::Success) {
FWGPU_LOGW << "Waiting for first frame work to finish resulted in an error"
<< static_cast<uint32_t>(wStatus);
}
firstRender = false;
}
mCommandBuffer = nullptr;
mTextureView = nullptr;
assert_invariant(mSwapChain);
@@ -1125,15 +1055,14 @@ void WebGPUDriver::bindPipeline(PipelineState const& pipelineState) {
assert_invariant(vertexBufferInfo);
std::vector<wgpu::TextureFormat> pipelineColorFormats;
wgpu::TextureFormat pipelineDepthStencilFormat = wgpu::TextureFormat::Undefined;
wgpu::TextureFormat pipelineDepthFormat = wgpu::TextureFormat::Undefined;
uint8_t pipelineSamples = 1;
bool const requestedDepth = any(mCurrentRenderTarget->getTargetFlags() & TargetBufferFlags::DEPTH);
bool const requestedStencil = any(mCurrentRenderTarget->getTargetFlags() & TargetBufferFlags::STENCIL);
pipelineSamples = mCurrentRenderTarget->getSamples();
if (mCurrentRenderTarget->isDefaultRenderTarget()) {
pipelineColorFormats.push_back(mSwapChain->getColorFormat());
pipelineDepthStencilFormat = mSwapChain->getDepthFormat();
pipelineDepthFormat = mSwapChain->getDepthFormat();
pipelineSamples =
mCurrentRenderTarget->getSamples();// Default RT should have samples (usually 1)
} else {
const auto& mrtColorAttachments = mCurrentRenderTarget->getColorAttachmentInfos();
for (size_t i = 0; i < MRT::MAX_SUPPORTED_RENDER_TARGET_COUNT; ++i) {
@@ -1147,19 +1076,21 @@ void WebGPUDriver::bindPipeline(PipelineState const& pipelineState) {
const auto& depthInfo = mCurrentRenderTarget->getDepthAttachmentInfo();
const auto& stencilInfo = mCurrentRenderTarget->getStencilAttachmentInfo();
Handle<HwTexture> depthStencilHandle = {};
if (depthInfo.handle) {
depthStencilHandle = depthInfo.handle;
} else if (stencilInfo.handle) {
depthStencilHandle = stencilInfo.handle;
}
if (depthStencilHandle) {
const auto dsTexture = handleCast<WebGPUTexture>(depthStencilHandle);
if (dsTexture) {
pipelineDepthStencilFormat = dsTexture->getTexture().GetFormat();
FILAMENT_CHECK_POSTCONDITION(!stencilInfo.handle)
<< "depth and stencil attachments cannot both be provided for WebGPU";
const auto depthTexture = handleCast<WebGPUTexture>(depthInfo.handle);
if (depthTexture) pipelineDepthFormat = depthTexture->getTexture().GetFormat();
} else {
if (stencilInfo.handle) {
const auto stencilTexture = handleCast<WebGPUTexture>(stencilInfo.handle);
// Assuming combined depth/stencil format if only stencil is present
if (stencilTexture) {
pipelineDepthFormat = stencilTexture->getTexture().GetFormat();
}
}
}
pipelineSamples = mCurrentRenderTarget->getSamples();
}
// TODO: We expected this to be a sane check, however it complains when running shadowtest.
@@ -1171,7 +1102,7 @@ void WebGPUDriver::bindPipeline(PipelineState const& pipelineState) {
wgpu::RenderPipeline pipeline = createWebGPURenderPipeline(mDevice, *program, *vertexBufferInfo,
layout, pipelineState.rasterState, pipelineState.stencilState,
pipelineState.polygonOffset, pipelineState.primitiveType, pipelineColorFormats,
pipelineDepthStencilFormat, pipelineSamples, requestedDepth, requestedStencil);
pipelineDepthFormat, pipelineSamples);
assert_invariant(pipeline);
mPipelineMap[hash] = pipeline;
mRenderPassEncoder.SetPipeline(pipeline);
+1 -7
View File
@@ -29,7 +29,6 @@
#include <utils/compiler.h>
#include "SpdMipmapGenerator/SpdMipmapGenerator.h"
#include <tsl/robin_map.h>
#include <webgpu/webgpu_cpp.h>
@@ -72,12 +71,12 @@ private:
WebGPUSwapChain* mSwapChain = nullptr;
uint64_t mNextFakeHandle = 1;
wgpu::CommandEncoder mCommandEncoder = nullptr;
std::vector<Handle<HwTexture>> mMipQueue;
wgpu::TextureView mTextureView = nullptr;
wgpu::RenderPassEncoder mRenderPassEncoder = nullptr;
wgpu::CommandBuffer mCommandBuffer = nullptr;
WebGPURenderTarget* mDefaultRenderTarget = nullptr;
WebGPURenderTarget* mCurrentRenderTarget = nullptr;
spd::MipmapGenerator mMipMapGenerator;
tsl::robin_map<uint32_t, wgpu::RenderPipeline> mPipelineMap;
@@ -121,11 +120,6 @@ private:
return mHandleAllocator.construct<D>(handle, std::forward<ARGS>(args)...);
}
template<typename D, typename B, typename... ARGS>
Handle<B> allocAndConstructHandle(ARGS&&... args) {
return mHandleAllocator.allocateAndConstruct<D>(std::forward<ARGS>(args)...);
}
template<typename D, typename B>
D* handleCast(Handle<B> handle) noexcept {
return mHandleAllocator.handle_cast<D*>(handle);
@@ -1,50 +0,0 @@
/*
* Copyright (C) 2025 The Android Open Source Project
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#include "WebGPUFence.h"
#include <backend/DriverEnums.h>
#include <webgpu/webgpu_cpp.h>
#include <atomic>
namespace filament::backend {
FenceStatus WebGPUFence::getStatus() { return mStatus.load(); }
void WebGPUFence::addMarkerToQueueState(wgpu::Queue const& queue) {
queue.OnSubmittedWorkDone(
wgpu::CallbackMode::AllowSpontaneous,
[](const wgpu::QueueWorkDoneStatus status,
std::atomic<FenceStatus>* const fenceStatus) {
if (!fenceStatus) {
return;
}
switch (status) {
case wgpu::QueueWorkDoneStatus::Success:
fenceStatus->store(FenceStatus::CONDITION_SATISFIED);
break;
case wgpu::QueueWorkDoneStatus::CallbackCancelled:
case wgpu::QueueWorkDoneStatus::Error:
fenceStatus->store(FenceStatus::ERROR);
break;
}
},
&mStatus);
}
} // namespace filament::backend
-43
View File
@@ -1,43 +0,0 @@
/*
* Copyright (C) 2025 The Android Open Source Project
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef TNT_FILAMENT_BACKEND_WEBGPUFENCE_H
#define TNT_FILAMENT_BACKEND_WEBGPUFENCE_H
#include "DriverBase.h"
#include <backend/DriverEnums.h>
#include <atomic>
namespace wgpu {
class Queue;
} // namespace wgpu
namespace filament::backend {
class WebGPUFence final : public HwFence {
public:
[[nodiscard]] FenceStatus getStatus();
void addMarkerToQueueState(wgpu::Queue const&);
private:
std::atomic<FenceStatus> mStatus{ FenceStatus::TIMEOUT_EXPIRED };
};
} // namespace filament::backend
#endif // TNT_FILAMENT_BACKEND_WEBGPUFENCE_H
@@ -169,52 +169,39 @@ wgpu::RenderPipeline createWebGPURenderPipeline(wgpu::Device const& device,
wgpu::PipelineLayout const& layout, RasterState const& rasterState,
StencilState const& stencilState, PolygonOffset const& polygonOffset,
const PrimitiveType primitiveType, std::vector<wgpu::TextureFormat> const& colorFormats,
wgpu::TextureFormat const& depthStencilFormat, const uint8_t samplesCount, const bool requestedDepth, const bool requestedStencil) {
const wgpu::TextureFormat depthFormat, const uint8_t samplesCount) {
assert_invariant(program.vertexShaderModule);
wgpu::DepthStencilState depthStencilState{};
const bool depthOrStencilRequested = (requestedDepth || requestedStencil);
if (depthFormat != wgpu::TextureFormat::Undefined) {
depthStencilState.format = depthFormat;
depthStencilState.depthWriteEnabled = rasterState.depthWrite;
depthStencilState.depthCompare = toWebGPU(rasterState.depthFunc);
depthStencilState.stencilFront = {
.compare = toWebGPU(stencilState.front.stencilFunc),
.failOp = toWebGPU(stencilState.front.stencilOpStencilFail),
.depthFailOp = toWebGPU(stencilState.front.stencilOpDepthFail),
.passOp = toWebGPU(stencilState.front.stencilOpDepthStencilPass),
};
depthStencilState.stencilBack = {
.compare = toWebGPU(stencilState.back.stencilFunc),
.failOp = toWebGPU(stencilState.back.stencilOpStencilFail),
.depthFailOp = toWebGPU(stencilState.back.stencilOpDepthFail),
.passOp = toWebGPU(stencilState.back.stencilOpDepthStencilPass),
};
depthStencilState.stencilReadMask =
stencilState.front.readMask; // Use front face's comparison mask for read mask
depthStencilState.stencilWriteMask = stencilState.stencilWrite ? 0xFFFFFFFF : 0u;
depthStencilState.depthBias = static_cast<int32_t>(polygonOffset.constant);
depthStencilState.depthBiasSlopeScale = polygonOffset.slope;
depthStencilState.depthBiasClamp = 0.0f;
if (depthOrStencilRequested) {
FILAMENT_CHECK_PRECONDITION(depthStencilFormat != wgpu::TextureFormat::Undefined)
<< "Depth or Stencil requested for pipeline, but depthStencilFormat is "
"wgpu::TextureFormat::Undefined.";
depthStencilState.format = depthStencilFormat;
if (requestedDepth) {
depthStencilState.depthWriteEnabled = rasterState.depthWrite;
depthStencilState.depthCompare = toWebGPU(rasterState.depthFunc);
depthStencilState.depthBias = static_cast<int32_t>(polygonOffset.constant);
depthStencilState.depthBiasSlopeScale = polygonOffset.slope;
depthStencilState.depthBiasClamp = 0.0f;
} else {
depthStencilState.depthWriteEnabled = false;
depthStencilState.depthCompare = wgpu::CompareFunction::Undefined;
depthStencilState.depthBias = 0;
depthStencilState.depthBiasSlopeScale = 0.0f;
depthStencilState.depthBiasClamp = 0.0f;
}
if (hasStencilAspect(depthStencilFormat) && requestedStencil) {
depthStencilState.stencilFront = {
.compare = toWebGPU(stencilState.front.stencilFunc),
.failOp = toWebGPU(stencilState.front.stencilOpStencilFail),
.depthFailOp = toWebGPU(stencilState.front.stencilOpDepthFail),
.passOp = toWebGPU(stencilState.front.stencilOpDepthStencilPass),
};
depthStencilState.stencilBack = {
.compare = toWebGPU(stencilState.back.stencilFunc),
.failOp = toWebGPU(stencilState.back.stencilOpStencilFail),
.depthFailOp = toWebGPU(stencilState.back.stencilOpDepthFail),
.passOp = toWebGPU(stencilState.back.stencilOpDepthStencilPass),
};
depthStencilState.stencilReadMask = stencilState.front.readMask;
depthStencilState.stencilWriteMask = stencilState.stencilWrite ? stencilState.front.writeMask : 0u;
} else {
depthStencilState.stencilFront.compare = wgpu::CompareFunction::Undefined;
if (!hasStencilAspect(depthFormat)) {
depthStencilState.stencilFront.compare = wgpu::CompareFunction::Always;
depthStencilState.stencilFront.failOp = wgpu::StencilOperation::Keep;
depthStencilState.stencilFront.depthFailOp = wgpu::StencilOperation::Keep;
depthStencilState.stencilFront.passOp = wgpu::StencilOperation::Keep;
depthStencilState.stencilBack = depthStencilState.stencilFront;
depthStencilState.stencilBack =
depthStencilState.stencilFront; // Keep back and front consistent
depthStencilState.stencilReadMask = 0;
depthStencilState.stencilWriteMask = 0;
}
@@ -258,11 +245,11 @@ wgpu::RenderPipeline createWebGPURenderPipeline(wgpu::Device const& device,
.unclippedDepth = !rasterState.depthClamp &&
device.HasFeature(wgpu::FeatureName::DepthClipControl)
},
.depthStencil = depthOrStencilRequested ? &depthStencilState: nullptr,
.depthStencil = depthFormat != wgpu::TextureFormat::Undefined ? &depthStencilState: nullptr,
.multisample = {
.count = samplesCount,
.mask = 0xFFFFFFFF,
.alphaToCoverageEnabled = (samplesCount > 1) && rasterState.alphaToCoverage
.alphaToCoverageEnabled = rasterState.alphaToCoverage
},
.fragment = nullptr // will add below if fragment module is included
};
@@ -40,8 +40,8 @@ class WebGPUProgram;
[[nodiscard]] wgpu::RenderPipeline createWebGPURenderPipeline(wgpu::Device const&,
WebGPUProgram const&, WebGPUVertexBufferInfo const&, wgpu::PipelineLayout const&,
RasterState const&, StencilState const&, PolygonOffset const&, PrimitiveType primitiveType,
std::vector<wgpu::TextureFormat> const& colorFormats, wgpu::TextureFormat const& depthStencilFormat,
uint8_t samplesCount, bool requestedDepth, bool requestedStencil);
std::vector<wgpu::TextureFormat> const& colorFormats, wgpu::TextureFormat depthFormat,
uint8_t samplesCount);
}// namespace filament::backend
@@ -32,27 +32,24 @@ namespace filament::backend {
WebGPURenderTarget::WebGPURenderTarget(const uint32_t width, const uint32_t height,
const uint8_t samples, const uint8_t layerCount, MRT const& colorAttachmentsMRT,
Attachment const& depthAttachmentInfo, Attachment const& stencilAttachmentInfo, TargetBufferFlags const& targetFlags)
Attachment const& depthAttachmentInfo, Attachment const& stencilAttachmentInfo)
: HwRenderTarget{ width, height },
mDefaultRenderTarget{ false },
mTargetFlags{ targetFlags },
mSamples{ samples },
mLayerCount{ layerCount },
mColorAttachments{ colorAttachmentsMRT },
mDepthAttachment{ depthAttachmentInfo },
mStencilAttachment{ stencilAttachmentInfo } {
// TODO consider making this an array
mColorAttachmentDesc.reserve(MRT::MAX_SUPPORTED_RENDER_TARGET_COUNT);
mColorAttachmentDescriptors.reserve(MRT::MAX_SUPPORTED_RENDER_TARGET_COUNT);
}
// Default constructor for the default render target
WebGPURenderTarget::WebGPURenderTarget()
: HwRenderTarget{ 0, 0 },
mDefaultRenderTarget{ true },
mTargetFlags{ TargetBufferFlags::NONE },
mSamples{ 1 },
mLayerCount{ 1 }
{}
mLayerCount{ 1 } {}
wgpu::LoadOp WebGPURenderTarget::getLoadOperation(RenderPassParams const& params,
const TargetBufferFlags bufferToOperateOn) {
@@ -60,7 +57,7 @@ wgpu::LoadOp WebGPURenderTarget::getLoadOperation(RenderPassParams const& params
return wgpu::LoadOp::Clear;
}
if (any(params.flags.discardStart & bufferToOperateOn)) {
return wgpu::LoadOp::Clear;
return wgpu::LoadOp::Clear; // Or wgpu::LoadOp::Undefined if clear is not desired on discard
}
return wgpu::LoadOp::Load;
}
@@ -76,30 +73,55 @@ wgpu::StoreOp WebGPURenderTarget::getStoreOperation(RenderPassParams const& para
void WebGPURenderTarget::setUpRenderPassAttachments(wgpu::RenderPassDescriptor& outDescriptor,
RenderPassParams const& params, wgpu::TextureView const& defaultColorTextureView,
wgpu::TextureView const& defaultDepthStencilTextureView,
wgpu::TextureFormat const& defaultDepthStencilFormat,
wgpu::TextureView const* customColorTextureViews, uint32_t customColorTextureViewCount,
wgpu::TextureView const& customDepthStencilTextureView){
mColorAttachmentDesc.clear();
wgpu::TextureView const& customDepthTextureView,
wgpu::TextureView const& customStencilTextureView,
const wgpu::TextureFormat customDepthFormat,
const wgpu::TextureFormat customStencilFormat) {
mColorAttachmentDescriptors.clear();
mHasDepthStencilAttachment = false;
const bool hasDepth = any(mTargetFlags & TargetBufferFlags::DEPTH);
const bool hasStencil = any(mTargetFlags & TargetBufferFlags::STENCIL);
mHasDepthStencilAttachment = hasDepth || hasStencil;
bool const depthReadOnly =
(params.readOnlyDepthStencil & RenderPassParams::READONLY_DEPTH) > 0;
// Color attachments
if (mDefaultRenderTarget) {
assert_invariant(defaultColorTextureView);
mColorAttachmentDesc.push_back({
.view = defaultColorTextureView,
mColorAttachmentDescriptors.push_back({ .view = defaultColorTextureView,
.resolveTarget = nullptr,
.loadOp = WebGPURenderTarget::getLoadOperation(params, TargetBufferFlags::COLOR0),
.storeOp = WebGPURenderTarget::getStoreOperation(params, TargetBufferFlags::COLOR0),
.clearValue = { params.clearColor.r, params.clearColor.g, params.clearColor.b,
params.clearColor.a } });
} else {
if (defaultDepthStencilTextureView) {
mDepthStencilAttachmentDescriptor = {
.view = defaultDepthStencilTextureView,
.depthLoadOp =
WebGPURenderTarget::getLoadOperation(params, TargetBufferFlags::DEPTH),
.depthStoreOp =
WebGPURenderTarget::getStoreOperation(params, TargetBufferFlags::DEPTH),
.depthClearValue = static_cast<float>(params.clearDepth),
.depthReadOnly =
(params.readOnlyDepthStencil & RenderPassParams::READONLY_DEPTH) > 0,
.stencilLoadOp = wgpu::LoadOp::Undefined,
.stencilStoreOp = wgpu::StoreOp::Undefined,
.stencilClearValue = params.clearStencil,
.stencilReadOnly =
(params.readOnlyDepthStencil & RenderPassParams::READONLY_STENCIL) > 0,
};
if (defaultDepthStencilFormat == wgpu::TextureFormat::Depth24PlusStencil8 ||
defaultDepthStencilFormat == wgpu::TextureFormat::Depth32FloatStencil8 ||
defaultDepthStencilFormat == wgpu::TextureFormat::Stencil8) {
mDepthStencilAttachmentDescriptor.stencilLoadOp =
WebGPURenderTarget::getLoadOperation(params, TargetBufferFlags::STENCIL);
mDepthStencilAttachmentDescriptor.stencilStoreOp =
WebGPURenderTarget::getStoreOperation(params, TargetBufferFlags::STENCIL);
}
mHasDepthStencilAttachment = true;
}
} else { // Custom Render Target
for (uint32_t i = 0; i < customColorTextureViewCount; ++i) {
if (customColorTextureViews[i]) {
mColorAttachmentDesc.push_back({ .view = customColorTextureViews[i],
mColorAttachmentDescriptors.push_back({ .view = customColorTextureViews[i],
// .resolveTarget = nullptr; // TODO: MSAA resolve for custom RT
.loadOp =
WebGPURenderTarget::getLoadOperation(params, getTargetBufferFlagsAt(i)),
@@ -111,62 +133,68 @@ void WebGPURenderTarget::setUpRenderPassAttachments(wgpu::RenderPassDescriptor&
.a = params.clearColor.a } });
}
}
}
outDescriptor.colorAttachmentCount = mColorAttachmentDesc.size();
outDescriptor.colorAttachments = mColorAttachmentDesc.data();
// Depth/Stencil Attachments
if (mHasDepthStencilAttachment) {
wgpu::TextureView depthStencilViewToUse{nullptr};
if (mDefaultRenderTarget) {
depthStencilViewToUse = defaultDepthStencilTextureView;
assert_invariant(depthStencilViewToUse);
} else {
if (customDepthStencilTextureView) {
depthStencilViewToUse = customDepthStencilTextureView;
}
}
FILAMENT_CHECK_POSTCONDITION(!(customDepthTextureView && customStencilTextureView))
<< "WebGPU CANNOT support separate texture views for depth + stencil. depth + "
"stencil needs to be in one texture view";
if (depthStencilViewToUse) {
mDepthStencilAttachmentDesc = {};
mDepthStencilAttachmentDesc.view = depthStencilViewToUse;
const bool hasStencil =
customStencilTextureView ||
(customDepthFormat == wgpu::TextureFormat::Depth24PlusStencil8 ||
customDepthFormat == wgpu::TextureFormat::Depth32FloatStencil8);
const bool hasDepth =
customDepthTextureView ||
(customStencilFormat == wgpu::TextureFormat::Depth24PlusStencil8 ||
customDepthFormat == wgpu::TextureFormat::Depth32FloatStencil8);
if (customDepthTextureView || customStencilTextureView) {
assert_invariant((hasDepth || hasStencil) &&
"Depth or Texture view without a valid texture format");
mDepthStencilAttachmentDescriptor = {};
mDepthStencilAttachmentDescriptor.view =
customDepthTextureView ? customDepthTextureView : customStencilTextureView;
if (hasDepth) {
mDepthStencilAttachmentDesc.depthLoadOp = depthReadOnly ? wgpu::LoadOp::Undefined :getLoadOperation(params, TargetBufferFlags::DEPTH);
mDepthStencilAttachmentDesc.depthStoreOp = depthReadOnly ? wgpu::StoreOp::Undefined :getStoreOperation(params, TargetBufferFlags::DEPTH);
mDepthStencilAttachmentDesc.depthClearValue = static_cast<float>(params.clearDepth);
mDepthStencilAttachmentDesc.depthReadOnly = depthReadOnly;
mDepthStencilAttachmentDescriptor.depthLoadOp =
WebGPURenderTarget::getLoadOperation(params, TargetBufferFlags::DEPTH);
mDepthStencilAttachmentDescriptor.depthStoreOp =
WebGPURenderTarget::getStoreOperation(params, TargetBufferFlags::DEPTH);
mDepthStencilAttachmentDescriptor.depthClearValue =
static_cast<float>(params.clearDepth);
mDepthStencilAttachmentDescriptor.depthReadOnly =
(params.readOnlyDepthStencil & RenderPassParams::READONLY_DEPTH) > 0;
} else {
mDepthStencilAttachmentDesc.depthLoadOp = wgpu::LoadOp::Undefined;
mDepthStencilAttachmentDesc.depthStoreOp = wgpu::StoreOp::Undefined;
mDepthStencilAttachmentDesc.depthReadOnly = true;
mDepthStencilAttachmentDescriptor.depthLoadOp = wgpu::LoadOp::Undefined;
mDepthStencilAttachmentDescriptor.depthStoreOp = wgpu::StoreOp::Undefined;
mDepthStencilAttachmentDescriptor.depthReadOnly = true;
}
if (hasStencil) {
mDepthStencilAttachmentDesc.stencilLoadOp = getLoadOperation(params, TargetBufferFlags::STENCIL);
mDepthStencilAttachmentDesc.stencilStoreOp = getStoreOperation(params, TargetBufferFlags::STENCIL);
mDepthStencilAttachmentDesc.stencilClearValue = params.clearStencil;
mDepthStencilAttachmentDesc.stencilReadOnly = (params.readOnlyDepthStencil & RenderPassParams::READONLY_STENCIL) > 0;
mDepthStencilAttachmentDescriptor.stencilLoadOp =
WebGPURenderTarget::getLoadOperation(params, TargetBufferFlags::STENCIL);
mDepthStencilAttachmentDescriptor.stencilStoreOp =
WebGPURenderTarget::getStoreOperation(params, TargetBufferFlags::STENCIL);
mDepthStencilAttachmentDescriptor.stencilClearValue = params.clearStencil;
mDepthStencilAttachmentDescriptor.stencilReadOnly =
(params.readOnlyDepthStencil & RenderPassParams::READONLY_STENCIL) > 0;
} else {
mDepthStencilAttachmentDesc.stencilLoadOp = wgpu::LoadOp::Undefined;
mDepthStencilAttachmentDesc.stencilStoreOp = wgpu::StoreOp::Undefined;
mDepthStencilAttachmentDesc.stencilReadOnly = true;
mDepthStencilAttachmentDescriptor.stencilLoadOp = wgpu::LoadOp::Undefined;
mDepthStencilAttachmentDescriptor.stencilStoreOp = wgpu::StoreOp::Undefined;
mDepthStencilAttachmentDescriptor.stencilReadOnly = true;
}
outDescriptor.depthStencilAttachment = &mDepthStencilAttachmentDesc;
} else {
outDescriptor.depthStencilAttachment = nullptr;
mDepthStencilAttachmentDesc = {};
mHasDepthStencilAttachment = true;
}
} else {
outDescriptor.depthStencilAttachment = nullptr;
mDepthStencilAttachmentDesc = {};
}
outDescriptor.colorAttachmentCount = mColorAttachmentDescriptors.size();
outDescriptor.colorAttachments = mColorAttachmentDescriptors.data();
outDescriptor.depthStencilAttachment =
mHasDepthStencilAttachment ? &mDepthStencilAttachmentDescriptor : nullptr;
// descriptor.sampleCount was removed from the core spec. If your webgpu.h still has it,
// and your Dawn version expects it, you might need to set it here based on this->samples.
// e.g., descriptor.sampleCount = this->samples;
}
} // namespace filament::backend
}// namespace filament::backend
@@ -34,41 +34,44 @@ public:
WebGPURenderTarget(uint32_t width, uint32_t height, uint8_t samples, uint8_t layerCount,
MRT const& colorAttachments, Attachment const& depthAttachment,
Attachment const& stencilAttachment, TargetBufferFlags const& targetFlags);
Attachment const& stencilAttachment);
// Default constructor for the default render target
WebGPURenderTarget();
// Updated signature: takes resolved views for custom RTs, and default views for default RT
void setUpRenderPassAttachments(
wgpu::RenderPassDescriptor& outDescriptor,
RenderPassParams const& params,
// For default render target:
wgpu::TextureView const& defaultColorTextureView,
wgpu::TextureView const& defaultDepthStencilTextureView,
wgpu::TextureFormat const& defaultDepthStencilFormat,
// For custom render targets:
wgpu::TextureView const* customColorTextureViews, // Array of views
uint32_t customColorTextureViewCount,
wgpu::TextureView const& customDepthStencilTextureView);
wgpu::TextureView const& customDepthTextureView,
wgpu::TextureView const& customStencilTextureView,
wgpu::TextureFormat customDepthFormat,
wgpu::TextureFormat customStencilFormat);
[[nodiscard]] bool isDefaultRenderTarget() const { return mDefaultRenderTarget; }
[[nodiscard]] uint8_t getSamples() const { return mSamples; }
[[nodiscard]] uint8_t getLayerCount() const { return mLayerCount; }
// Accessors for the driver to get stored attachment info
[[nodiscard]] MRT const& getColorAttachmentInfos() const { return mColorAttachments; }
[[nodiscard]] Attachment const& getDepthAttachmentInfo() const { return mDepthAttachment; }
[[nodiscard]] Attachment const& getStencilAttachmentInfo() const { return mStencilAttachment; }
// Static helpers for load/store operations
[[nodiscard]] static wgpu::LoadOp getLoadOperation(RenderPassParams const& params,
TargetBufferFlags buffer);
[[nodiscard]] static wgpu::StoreOp getStoreOperation(RenderPassParams const& params,
TargetBufferFlags buffer);
[[nodiscard]] TargetBufferFlags getTargetFlags() const { return mTargetFlags; }
void setTargetFlags( TargetBufferFlags value) { mTargetFlags = value; }
private:
bool mDefaultRenderTarget = false;
TargetBufferFlags mTargetFlags = TargetBufferFlags::NONE;
uint8_t mSamples = 1;
uint8_t mLayerCount = 1;
@@ -79,8 +82,8 @@ private:
Attachment mStencilAttachment{};
// Cached descriptors for the render pass
std::vector<wgpu::RenderPassColorAttachment> mColorAttachmentDesc;
wgpu::RenderPassDepthStencilAttachment mDepthStencilAttachmentDesc{};
std::vector<wgpu::RenderPassColorAttachment> mColorAttachmentDescriptors;
wgpu::RenderPassDepthStencilAttachment mDepthStencilAttachmentDescriptor{};
bool mHasDepthStencilAttachment = false;
};
+66 -14
View File
@@ -21,27 +21,50 @@
#include <backend/DriverEnums.h>
#include <utils/Panic.h>
#include <utils/ostream.h>
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
#include <dawn/webgpu_cpp_print.h>
#endif
#include <webgpu/webgpu_cpp.h>
#include <cstdint>
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
#include <sstream>
#include <string>
#endif
#include <string_view>
/**
* Reusable set of convenience functions for strings -- generally string views, literals, &
* strings -- used in the WebGPU backend
* streaming -- used in the WebGPU backend
*/
namespace filament::backend {
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
/**
* Convenience template to print WGPU types (which don't support Filament's utils::io::ostream)
* @tparam WebGPUPrintable e.g. wgpu::FeatureName type
* @param out stream to print to
* @param printable the instance to print
* @return the stream printed to
*
* Example usage (because including the template operator override itself does not compile):
* #if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
* utils::io::ostream& operator<<(utils::io::ostream& out,
* const wgpu::FeatureName featureName) noexcept {
* return streamInsertWebGPUPrintable(out, featureName);
* }
* #endif
*/
template<typename WebGPUPrintable>
[[nodiscard]] inline std::string webGPUPrintableToString(const WebGPUPrintable printable) {
std::stringstream out;
out << printable;
return out.str();
utils::io::ostream& streamInsertWebGPUPrintable(utils::io::ostream& out,
const WebGPUPrintable printable) {
std::stringstream printableStream;
printableStream << printable;
out << printableStream.str();
return out;
}
#endif
@@ -92,17 +115,46 @@ template<typename WebGPUPrintable>
}
}
[[nodiscard]] inline std::string adapterOptionsToString(
wgpu::RequestAdapterOptions const& options) {
std::stringstream out;
/**
* Convenience template to print wgpu::RequestAdapterOptions
* @tparam STREAM_TYPE useful namely for streaming to either an utils::io::ostream OR
* utils::details::PanicStream, but also for any kind of stream
* @param out stream to print to
* @param options instance to be printed
* @return stream printed to
*
* Example usage (because including the template operator override itself does not compile):
* template<typename STREAM_TYPE>
* STREAM_TYPE& operator<<(STREAM_TYPE& out, wgpu::RequestAdapterOptions const& options) noexcept {
* return streamInsertRequestAdapterOptions(out, options);
* }
*/
template<typename STREAM_TYPE>
STREAM_TYPE& streamInsertRequestAdapterOptions(STREAM_TYPE& out,
wgpu::RequestAdapterOptions const& options) noexcept {
out << "power preference " << powerPreferenceToString(options.powerPreference)
<< " force fallback adapter " << bool(options.forceFallbackAdapter) << " backend type "
<< backendTypeToString(options.backendType);
return out.str();
return out;
}
[[nodiscard]] inline std::string adapterInfoToString(wgpu::AdapterInfo const& info) {
std::stringstream out;
/**
* Convenience template to print wgpu::AdapterInfo
* @tparam STREAM_TYPE useful namely for streaming to either an utils::io::ostream OR
* utils::details::PanicStream, but also for any kind of stream
* @param out stream to print to
* @param options instance to be printed
* @return stream printed to
*
* Example usage (because including the template operator override itself does not compile):
* template<typename STREAM_TYPE>
* STREAM_TYPE& operator<<(STREAM_TYPE& out, wgpu::AdapterInfo const& info) noexcept {
* return streamInsertRequestAdapterInfo(out, info);
* }
*/
template<class STREAM_TYPE>
STREAM_TYPE& streamInsertRequestAdapterInfo(STREAM_TYPE& out,
wgpu::AdapterInfo const& info) noexcept {
out << "vendor (" << info.vendorID << ") '" << info.vendor
<< "' device (" << info.deviceID << ") '" << info.device
<< "' adapter " << adapterTypeToString(info.adapterType)
@@ -110,7 +162,7 @@ template<typename WebGPUPrintable>
<< " architecture '" << info.architecture
<< "' subgroupMinSize " << info.subgroupMinSize
<< " subgroupMaxSize " << info.subgroupMaxSize;
return out.str();
return out;
}
[[nodiscard]] constexpr std::string_view deviceLostReasonToString(
@@ -123,7 +175,7 @@ template<typename WebGPUPrintable>
}
}
[[nodiscard]] constexpr std::string_view filamentShaderStageToString(const ShaderStage stage) {
[[nodiscard]] constexpr std::string_view filamentShaderStageToString(ShaderStage stage) {
switch (stage) {
case ShaderStage::VERTEX: return "vertex";
case ShaderStage::FRAGMENT: return "fragment";
+35 -23
View File
@@ -24,25 +24,43 @@
#include "backend/DriverEnums.h"
#include <utils/Panic.h>
#include <utils/ostream.h>
#include <webgpu/webgpu_cpp.h>
#include <algorithm>
#include <cstdint>
namespace filament::backend {
namespace {
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
utils::io::ostream& operator<<(utils::io::ostream& out, const wgpu::TextureFormat format) noexcept {
return filament::backend::streamInsertWebGPUPrintable(out, format);
}
utils::io::ostream& operator<<(utils::io::ostream& out,
const wgpu::TextureUsage textureUsage) noexcept {
return filament::backend::streamInsertWebGPUPrintable(out, textureUsage);
}
utils::io::ostream& operator<<(utils::io::ostream& out,
const wgpu::PresentMode presentMode) noexcept {
return filament::backend::streamInsertWebGPUPrintable(out, presentMode);
}
utils::io::ostream& operator<<(utils::io::ostream& out,
const wgpu::CompositeAlphaMode alphaMode) noexcept {
return filament::backend::streamInsertWebGPUPrintable(out, alphaMode);
}
void printSurfaceCapabilitiesDetails(wgpu::SurfaceCapabilities const& capabilities) {
FWGPU_LOGI << "WebGPU surface capabilities:";
FWGPU_LOGI << " surface usages: " << webGPUPrintableToString(capabilities.usages);
FWGPU_LOGI << " surface usages: " << capabilities.usages;
FWGPU_LOGI << " surface formats (" << capabilities.formatCount << "):";
if (capabilities.formatCount > 0 && capabilities.formats != nullptr) {
std::for_each(capabilities.formats, capabilities.formats + capabilities.formatCount,
[](wgpu::TextureFormat const format) {
FWGPU_LOGI << " " << webGPUPrintableToString(format);
FWGPU_LOGI << " " << format;
});
}
FWGPU_LOGI << " surface present modes (" << capabilities.presentModeCount << "):";
@@ -50,7 +68,7 @@ void printSurfaceCapabilitiesDetails(wgpu::SurfaceCapabilities const& capabiliti
std::for_each(capabilities.presentModes,
capabilities.presentModes + capabilities.presentModeCount,
[](wgpu::PresentMode const presentMode) {
FWGPU_LOGI << " " << webGPUPrintableToString(presentMode);
FWGPU_LOGI << " " << presentMode;
});
}
FWGPU_LOGI << " surface alpha modes (" << capabilities.alphaModeCount << "):";
@@ -58,7 +76,7 @@ void printSurfaceCapabilitiesDetails(wgpu::SurfaceCapabilities const& capabiliti
std::for_each(capabilities.alphaModes,
capabilities.alphaModes + capabilities.alphaModeCount,
[](wgpu::CompositeAlphaMode const alphaMode) {
FWGPU_LOGI << " " << webGPUPrintableToString(alphaMode);
FWGPU_LOGI << " " << alphaMode;
});
}
}
@@ -66,20 +84,20 @@ void printSurfaceCapabilitiesDetails(wgpu::SurfaceCapabilities const& capabiliti
void printSurfaceConfiguration(wgpu::SurfaceConfiguration const& config,
wgpu::TextureFormat depthFormat) {
FWGPU_LOGI << "WebGPU surface configuration:";
FWGPU_LOGI << " surface format: " << webGPUPrintableToString(config.format);
FWGPU_LOGI << " surface usage: " << webGPUPrintableToString(config.usage);
FWGPU_LOGI << " surface format: " << config.format;
FWGPU_LOGI << " surface usage: " << config.usage;
FWGPU_LOGI << " surface view formats (" << config.viewFormatCount << "):";
if (config.viewFormatCount > 0 && config.viewFormats != nullptr) {
std::for_each(config.viewFormats, config.viewFormats + config.viewFormatCount,
[](wgpu::TextureFormat const viewFormat) {
FWGPU_LOGI << " " << webGPUPrintableToString(viewFormat);
FWGPU_LOGI << " " << viewFormat;
});
}
FWGPU_LOGI << " surface alpha mode: " << webGPUPrintableToString(config.alphaMode);
FWGPU_LOGI << " surface alpha mode: " << config.alphaMode;
FWGPU_LOGI << " surface width: " << config.width;
FWGPU_LOGI << " surface height: " << config.height;
FWGPU_LOGI << " surface present mode: " << webGPUPrintableToString(config.presentMode);
FWGPU_LOGI << "WebGPU selected depth format: " << webGPUPrintableToString(depthFormat);
FWGPU_LOGI << " surface present mode: " << config.presentMode;
FWGPU_LOGI << "WebGPU selected depth format: " << depthFormat;
}
#endif// FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
@@ -110,6 +128,7 @@ void printSurfaceConfiguration(wgpu::SurfaceConfiguration const& config,
return wgpu::TextureFormat::Depth24PlusStencil8;
}
} else {
// other options: Depth16Unorm or Depth24Plus
return wgpu::TextureFormat::Depth32Float;
}
}
@@ -216,9 +235,10 @@ void initConfig(wgpu::SurfaceConfiguration& config, wgpu::Device const& device,
}
[[nodiscard]] wgpu::TextureView createDepthTextureView(wgpu::Texture const& depthTexture,
wgpu::TextureFormat const& depthFormat, bool const needStencil) {
wgpu::TextureFormat depthFormat, bool needStencil) {
wgpu::TextureViewDescriptor descriptor{
.label = "depth_texture_view",
.format = depthFormat,
.dimension = wgpu::TextureViewDimension::e2D,
.baseMipLevel = 0,
.mipLevelCount = 1,
@@ -229,16 +249,6 @@ void initConfig(wgpu::SurfaceConfiguration& config, wgpu::Device const& device,
};
if (needStencil) {
descriptor.aspect = wgpu::TextureAspect::All;
descriptor.format = depthFormat;
} else {
descriptor.aspect = wgpu::TextureAspect::DepthOnly;
if (depthFormat == wgpu::TextureFormat::Depth32FloatStencil8) {
descriptor.format = wgpu::TextureFormat::Depth32Float;
} else if (depthFormat == wgpu::TextureFormat::Depth24PlusStencil8) {
descriptor.format = wgpu::TextureFormat::Depth24Plus;
} else {
descriptor.format = depthFormat;
}
}
wgpu::TextureView depthTextureView = depthTexture.CreateView(&descriptor);
FILAMENT_CHECK_POSTCONDITION(depthTextureView) << "Failed to create depth texture view";
@@ -247,6 +257,8 @@ void initConfig(wgpu::SurfaceConfiguration& config, wgpu::Device const& device,
}// namespace
namespace filament::backend {
WebGPUSwapChain::WebGPUSwapChain(wgpu::Surface&& surface, wgpu::Extent2D const& surfaceSize,
wgpu::Adapter const& adapter, wgpu::Device const& device, uint64_t flags)
: mDevice(device),
+32 -16
View File
@@ -25,6 +25,7 @@
#include <utils/BitmaskEnum.h>
#include <utils/Panic.h>
#include <utils/debug.h>
#include <utils/ostream.h>
#include <webgpu/webgpu_cpp.h>
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
@@ -342,7 +343,8 @@ WebGPUTexture::WebGPUTexture(const SamplerType samplerType, const uint8_t levels
std::stringstream textureFormatStream;
textureFormatStream << mWebGPUFormat;
FWGPU_LOGD << "Texture '" << textureDescriptor.label << "' has view format "
<< viewFormatStream.str() << " and texture format " << textureFormatStream.str();
<< viewFormatStream.str() << " and texture format " << textureFormatStream.str()
<< utils::io::endl;
}
#endif
}
@@ -468,43 +470,51 @@ wgpu::TextureFormat WebGPUTexture::fToWGPUTextureFormat(TextureFormat const& fFo
// No direct mapping in wgpu. Could potentially map to RGBA8Unorm
// and discard the alpha and lower precision.
FWGPU_LOGW << "Requested Filament texture format RGB565 but getting "
"wgpu::TextureFormat::Undefined (no direct mapping in wgpu)";
"wgpu::TextureFormat::Undefined (no direct mapping in wgpu)"
<< utils::io::endl;
return wgpu::TextureFormat::Undefined;
case TextureFormat::RGB9_E5: return wgpu::TextureFormat::RGB9E5Ufloat;
case TextureFormat::RGB5_A1:
// No direct mapping in wgpu. Could potentially map to RGBA8Unorm
// and handle the packing/unpacking in shaders.
FWGPU_LOGW << "Requested Filament texture format RGB5_A1 but getting "
"wgpu::TextureFormat::Undefined (no direct mapping in wgpu)";
"wgpu::TextureFormat::Undefined (no direct mapping in wgpu)"
<< utils::io::endl;
return wgpu::TextureFormat::Undefined;
case TextureFormat::RGBA4:
// No direct mapping in wgpu. Could potentially map to RGBA8Unorm
// and handle the packing/unpacking in shaders.
FWGPU_LOGW << "Requested Filament texture format RGBA4 but getting "
"wgpu::TextureFormat::Undefined (no direct mapping in wgpu)";
"wgpu::TextureFormat::Undefined (no direct mapping in wgpu)"
<< utils::io::endl;
return wgpu::TextureFormat::Undefined;
case TextureFormat::RGB8:
FWGPU_LOGW << "Requested Filament texture format RGB8 but getting "
"wgpu::TextureFormat::RGBA8Unorm (no direct sRGB equivalent in wgpu "
"without alpha)";
"without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA8Unorm;
case TextureFormat::SRGB8:
FWGPU_LOGW << "Requested Filament texture format SRGB8 but getting "
"wgpu::TextureFormat::RGBA8UnormSrgb (no direct sRGB equivalent in wgpu "
"without alpha)";
"without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA8UnormSrgb;
case TextureFormat::RGB8_SNORM:
FWGPU_LOGW
<< "Requested Filament texture format RGB8_SNORM but getting "
"wgpu::TextureFormat::RGBA8Snorm (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA8Snorm (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA8Snorm;
case TextureFormat::RGB8UI:
FWGPU_LOGW << "Requested Filament texture format RGB8UI but getting "
"wgpu::TextureFormat::RGBA8Uint (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA8Uint (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA8Uint;
case TextureFormat::RGB8I:
FWGPU_LOGW << "Requested Filament texture format RGB8I but getting "
"wgpu::TextureFormat::RGBA8Sint (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA8Sint (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA8Sint;
case TextureFormat::R11F_G11F_B10F: return wgpu::TextureFormat::RG11B10Ufloat;
case TextureFormat::UNUSED: return wgpu::TextureFormat::Undefined;
@@ -512,32 +522,38 @@ wgpu::TextureFormat WebGPUTexture::fToWGPUTextureFormat(TextureFormat const& fFo
case TextureFormat::RGB16F:
FWGPU_LOGW
<< "Requested Filament texture format RGB16F but getting "
"wgpu::TextureFormat::RGBA16Float (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA16Float (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA16Float;
case TextureFormat::RGB16UI:
FWGPU_LOGW
<< "Requested Filament texture format RGB16UI but getting "
"wgpu::TextureFormat::RGBA16Uint (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA16Uint (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA16Uint;
case TextureFormat::RGB16I:
FWGPU_LOGW
<< "Requested Filament texture format RGB16I but getting "
"wgpu::TextureFormat::RGBA16Sint (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA16Sint (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA16Sint;
case TextureFormat::RGB32F:
FWGPU_LOGW
<< "Requested Filament texture format RGB32F but getting "
"wgpu::TextureFormat::RGBA32Float (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA32Float (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA32Float;
case TextureFormat::RGB32UI:
FWGPU_LOGW
<< "Requested Filament texture format RGB32UI but getting "
"wgpu::TextureFormat::RGBA32Uint (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA32Uint (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA32Uint;
case TextureFormat::RGB32I:
FWGPU_LOGW
<< "Requested Filament texture format RGB32I but getting "
"wgpu::TextureFormat::RGBA32Sint (no direct mapping in wgpu without alpha)";
"wgpu::TextureFormat::RGBA32Sint (no direct mapping in wgpu without alpha)"
<< utils::io::endl;
return wgpu::TextureFormat::RGBA32Sint;
case TextureFormat::DXT1_RGB: return wgpu::TextureFormat::BC1RGBAUnorm;
case TextureFormat::DXT1_RGBA: return wgpu::TextureFormat::BC1RGBAUnorm;
@@ -557,7 +573,7 @@ wgpu::TextureView WebGPUTexture::makeTextureView(const uint8_t& baseLevel,
if (baseLevel > 0 && !mSupportsMultipleMipLevels) {
FWGPU_LOGW << "Trying to make a texture view into a level ("
<< static_cast<uint32_t>(baseLevel)
<< ") for which we cannot generate mip levels.";
<< ") for which we cannot generate mip levels." << utils::io::endl;
}
#endif
const wgpu::TextureViewDescriptor textureViewDescriptor{
@@ -22,6 +22,7 @@
#include <backend/DriverEnums.h>
#include <utils/Panic.h>
#include <utils/ostream.h>
#include <webgpu/webgpu_cpp.h>
@@ -56,22 +57,26 @@ namespace {
case ElementType::BYTE3:
FWGPU_LOGW
<< "Requested Filament vertex format BYTE3 (normalized) but getting "
"wgpu::VertexFormat::Snorm8x4 (no direct mapping in wgpu for x3 byte)";
"wgpu::VertexFormat::Snorm8x4 (no direct mapping in wgpu for x3 byte)"
<< utils::io::endl;
return VertexFormat::Snorm8x4; // NOT MINSPEC
case ElementType::UBYTE3:
FWGPU_LOGW
<< "Requested Filament vertex format UBYTE3 (normalized) but getting "
"wgpu::VertexFormat::Unorm8x4 (no direct mapping in wgpu for x3 byte)";
"wgpu::VertexFormat::Unorm8x4 (no direct mapping in wgpu for x3 byte)"
<< utils::io::endl;
return VertexFormat::Unorm8x4; // NOT MINSPEC
case ElementType::SHORT3:
FWGPU_LOGW << "Requested Filament vertex format SHORT3 (normalized) but getting "
"wgpu::VertexFormat::Snorm16x4 (no direct mapping in wgpu for x3 "
"half/short)";
"half/short)"
<< utils::io::endl;
return VertexFormat::Snorm16x4; // NOT MINSPEC
case ElementType::USHORT3:
FWGPU_LOGW << "Requested Filament vertex format USHORT3 (normalized) but getting "
"wgpu::VertexFormat::Unorm16x4 (no direct mapping in wgpu for x3 "
"half/short)";
"half/short)"
<< utils::io::endl;
return VertexFormat::Unorm16x4; // NOT MINSPEC
// Four Component Types
case ElementType::BYTE4: return VertexFormat::Snorm8x4;
@@ -92,22 +97,26 @@ namespace {
case ElementType::BYTE:
if (integer) return VertexFormat::Sint8;
FWGPU_LOGW << "Requested Filament vertex format BYTE (float) but getting "
"wgpu::VertexFormat::Float16 (no direct mapping in wgpu for 8 bit float)";
"wgpu::VertexFormat::Float16 (no direct mapping in wgpu for 8 bit float)"
<< utils::io::endl;
return wgpu::VertexFormat::Float16;
case ElementType::UBYTE:
if (integer) return VertexFormat::Uint8;
FWGPU_LOGW << "Requested Filament vertex format UBYTE (float) but getting "
"wgpu::VertexFormat::Float16 (no direct mapping in wgpu for 8 bit float)";
"wgpu::VertexFormat::Float16 (no direct mapping in wgpu for 8 bit float)"
<< utils::io::endl;
return VertexFormat::Float16;
case ElementType::SHORT:
if (integer) return VertexFormat::Sint16;
FWGPU_LOGW << "Requested Filament vertex format SHORT (float) and getting "
"wgpu::VertexFormat::Float16 (potential loss of precision?)";
"wgpu::VertexFormat::Float16 (potential loss of precision?)"
<< utils::io::endl;
return VertexFormat ::Float16;
case ElementType::USHORT:
if (integer) return VertexFormat::Uint16;
FWGPU_LOGW << "Requested Filament vertex format USHORT (float) and getting "
"wgpu::VertexFormat::Float16 (potential loss of precision?)";
"wgpu::VertexFormat::Float16 (potential loss of precision?)"
<< utils::io::endl;
return VertexFormat::Float16;
case ElementType::HALF: return VertexFormat::Float16;
case ElementType::INT: return VertexFormat::Sint32;
@@ -118,49 +127,58 @@ namespace {
if (integer) return VertexFormat::Sint8x2;
FWGPU_LOGW
<< "Requested Filament vertex format BYTE2 (float) but getting "
"wgpu::VertexFormat::Float16x2 (no direct mapping in wgpu for 8 bit float)";
"wgpu::VertexFormat::Float16x2 (no direct mapping in wgpu for 8 bit float)"
<< utils::io::endl;
return VertexFormat::Float16x2;
case ElementType::UBYTE2:
if (integer) return VertexFormat::Uint8x2;
FWGPU_LOGW
<< "Requested Filament vertex format UBYTE2 (float) but getting "
"wgpu::VertexFormat::Float16x2 (no direct mapping in wgpu for 8 bit float)";
"wgpu::VertexFormat::Float16x2 (no direct mapping in wgpu for 8 bit float)"
<< utils::io::endl;
return VertexFormat::Float16x2;
case ElementType::SHORT2:
if (integer) return VertexFormat::Sint16x2;
FWGPU_LOGW << "Requested Filament vertex format SHORT2 (float) but getting "
"wgpu::VertexFormat::Float16x2 (potential loss of precision?)";
"wgpu::VertexFormat::Float16x2 (potential loss of precision?)"
<< utils::io::endl;
return VertexFormat::Float16x2;
case ElementType::USHORT2:
if (integer) return VertexFormat::Uint16x2;
FWGPU_LOGW << "Requested Filament vertex format USHORT2 (float) but getting "
"wgpu::VertexFormat::Float16x2 (potential loss of precision?)";
"wgpu::VertexFormat::Float16x2 (potential loss of precision?)"
<< utils::io::endl;
return VertexFormat::Float16x2;
case ElementType::HALF2: return VertexFormat::Float16x2;
case ElementType::FLOAT2: return VertexFormat::Float32x2;
// Three Component Types
case ElementType::BYTE3:
FWGPU_LOGW << "Requested Filament vertex format BYTE3 but getting "
"wgpu::VertexFormat::Sint8x4 (no direct mapping in wgpu for x3 byte)";
"wgpu::VertexFormat::Sint8x4 (no direct mapping in wgpu for x3 byte)"
<< utils::io::endl;
return VertexFormat::Sint8x4; // NOT MINSPEC
case ElementType::UBYTE3:
FWGPU_LOGW << "Requested Filament vertex format UBYTE3 but getting "
"wgpu::VertexFormat::Uint8x4 (no direct mapping in wgpu for x3 byte)";
"wgpu::VertexFormat::Uint8x4 (no direct mapping in wgpu for x3 byte)"
<< utils::io::endl;
return VertexFormat::Uint8x4; // NOT MINSPEC
case ElementType::SHORT3:
FWGPU_LOGW
<< "Requested Filament vertex format SHORT3 but getting "
"wgpu::VertexFormat::Sint16x4 (no direct mapping in wgpu for x3 half/short)";
"wgpu::VertexFormat::Sint16x4 (no direct mapping in wgpu for x3 half/short)"
<< utils::io::endl;
return VertexFormat::Sint16x4; // NOT MINSPEC
case ElementType::USHORT3:
FWGPU_LOGW
<< "Requested Filament vertex format USHORT3 but getting "
"wgpu::VertexFormat::Uint16x4 (no direct mapping in wgpu for x3 half/short)";
"wgpu::VertexFormat::Uint16x4 (no direct mapping in wgpu for x3 half/short)"
<< utils::io::endl;
return VertexFormat::Uint16x4; // NOT MINSPEC
case ElementType::HALF3:
FWGPU_LOGW
<< "Requested Filament vertex format HALF3 but getting "
"wgpu::VertexFormat::Float16x4 (no direct mapping in wgpu for x3 half/short)";
"wgpu::VertexFormat::Float16x4 (no direct mapping in wgpu for x3 half/short)"
<< utils::io::endl;
return VertexFormat::Float16x4; // NOT MINSPEC
case ElementType::FLOAT3: return VertexFormat::Float32x3;
// Four Component Types
@@ -168,23 +186,27 @@ namespace {
if (integer) return VertexFormat::Sint8x4;
FWGPU_LOGW
<< "Requested Filament vertex format BYTE4 (float) but getting "
"wgpu::VertexFormat::Float16x4 (no direct mapping in wgpu for 8 bit float)";
"wgpu::VertexFormat::Float16x4 (no direct mapping in wgpu for 8 bit float)"
<< utils::io::endl;
return VertexFormat::Float16x4;
case ElementType::UBYTE4:
if (integer) return VertexFormat::Uint8x4;
FWGPU_LOGW
<< "Requested Filament vertex format UBYTE4 (float) but getting "
"wgpu::VertexFormat::Float16x4 (no direct mapping in wgpu for 8 bit float)";
"wgpu::VertexFormat::Float16x4 (no direct mapping in wgpu for 8 bit float)"
<< utils::io::endl;
return VertexFormat::Float16x4;
case ElementType::SHORT4:
if (integer) return VertexFormat::Sint16x4;
FWGPU_LOGW << "Requested Filament vertex format SHORT4 (float) but getting "
"wgpu::VertexFormat::Float16x4 (potential loss of precision?)";
"wgpu::VertexFormat::Float16x4 (potential loss of precision?)"
<< utils::io::endl;
return VertexFormat::Float16x4;
case ElementType::USHORT4:
if (integer) return VertexFormat::Uint16x4;
FWGPU_LOGW << "Requested Filament vertex format USHORT4 (float) but getting "
"wgpu::VertexFormat::Float16x4 (potential loss of precision?)";
"wgpu::VertexFormat::Float16x4 (potential loss of precision?)"
<< utils::io::endl;
return VertexFormat::Float16x4;
case ElementType::HALF4: return VertexFormat::Float16x4;
case ElementType::FLOAT4: return VertexFormat::Float32x4;
@@ -24,6 +24,7 @@
#include <backend/Platform.h>
#include <utils/Hash.h>
#include <utils/Panic.h>
#include <utils/ostream.h>
#include <dawn/webgpu_cpp_print.h>
#include <webgpu/webgpu_cpp.h>
@@ -38,7 +39,6 @@
#include <utility>
#include <vector>
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
#include <string>
#include <string_view>
#include <variant>
#endif
@@ -53,9 +53,6 @@ namespace filament::backend {
namespace {
// The SPD Algorithm can make use of up to 12 storage texture attachments
constexpr uint32_t MAX_MIPMAP_STORAGE_TEXTURES_PER_STAGE = 12u;
constexpr std::array REQUIRED_FEATURES = { wgpu::FeatureName::TransientAttachments,
/*To make filtering assumptions like we want while waiting for Filament to provide that info,
float 32 needs to be filterable*/
@@ -120,6 +117,28 @@ constexpr void forEachLimitToValidate(std::function<bool(LimitToValidate const)>
}
}
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
utils::io::ostream& operator<<(utils::io::ostream& out,
const wgpu::WGSLLanguageFeatureName languageFeatureName) noexcept {
return streamInsertWebGPUPrintable(out, languageFeatureName);
}
utils::io::ostream& operator<<(utils::io::ostream& out,
const wgpu::FeatureName featureName) noexcept {
return streamInsertWebGPUPrintable(out, featureName);
}
#endif
template<typename STREAM_TYPE>
STREAM_TYPE& operator<<(STREAM_TYPE& out, wgpu::RequestAdapterOptions const& options) noexcept {
return streamInsertRequestAdapterOptions(out, options);
}
template<class STREAM_TYPE>
STREAM_TYPE& operator<<(STREAM_TYPE& out, wgpu::AdapterInfo const& info) noexcept {
return streamInsertRequestAdapterInfo(out, info);
}
constexpr bool isDefined(uint32_t limit) { return limit < wgpu::kLimitU32Undefined; }
[[maybe_unused]] constexpr bool isDefined(uint64_t limit) {
@@ -240,7 +259,7 @@ void printInstanceDetails(wgpu::Instance const& instance) {
supportedWGSLLanguageFeatures.features +
supportedWGSLLanguageFeatures.featureCount,
[](wgpu::WGSLLanguageFeatureName const featureName) {
FWGPU_LOGI << " " << webGPUPrintableToString(featureName);
FWGPU_LOGI << " " << featureName;
});
}
}
@@ -280,22 +299,23 @@ void printInstanceDetails(wgpu::Instance const& instance) {
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
void printLimit(std::string_view name, const std::variant<uint32_t, uint64_t> value) {
static constexpr std::string_view indent = " ";
FWGPU_LOGI << " " << name.data() << ": ";
bool undefined = true;
if (std::holds_alternative<uint32_t>(value)) {
if (std::get<uint32_t>(value) != WGPU_LIMIT_U32_UNDEFINED) {
undefined = false;
FWGPU_LOGI << indent << name.data() << ": " << std::get<uint32_t>(value);
FWGPU_LOGI << std::get<uint32_t>(value);
}
} else if (std::holds_alternative<uint64_t>(value)) {
if (std::get<uint64_t>(value) != WGPU_LIMIT_U64_UNDEFINED) {
undefined = false;
FWGPU_LOGI << indent << name.data() << ": " << std::get<uint64_t>(value);
FWGPU_LOGI << std::get<uint64_t>(value);
}
}
if (undefined) {
FWGPU_LOGI << indent << name.data() << ": UNDEFINED";
FWGPU_LOGI << "UNDEFINED";
}
FWGPU_LOGI << "";
}
#endif// FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
@@ -373,16 +393,16 @@ struct AdapterDetails final {
}
};
[[nodiscard]] std::string toString(AdapterDetails const& details) {
std::stringstream out;
out << adapterInfoToString(details.info)
template<class STREAM_TYPE>
STREAM_TYPE& operator<<(STREAM_TYPE& out, AdapterDetails const& details) noexcept {
out << details.info
<< " power preference " << powerPreferenceToString(details.powerPreference);
return out.str();
return out;
}
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
void printAdapterDetails(AdapterDetails const& details) {
FWGPU_LOGI << "Selected WebGPU adapter info: " << toString(details);
FWGPU_LOGI << "Selected WebGPU adapter info: " << details;
wgpu::SupportedFeatures supportedFeatures{};
details.adapter.GetFeatures(&supportedFeatures);
FWGPU_LOGI << "WebGPU adapter supported features (" << supportedFeatures.featureCount
@@ -391,7 +411,7 @@ void printAdapterDetails(AdapterDetails const& details) {
std::for_each(supportedFeatures.features,
supportedFeatures.features + supportedFeatures.featureCount,
[](wgpu::FeatureName const featureName) {
FWGPU_LOGI << " " << webGPUPrintableToString(featureName);
FWGPU_LOGI << " " << featureName;
});
}
wgpu::Limits supportedLimits{};
@@ -424,12 +444,12 @@ struct AdapterDetailsHash final {
return details.adapter.HasFeature(featureName);
})) {
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
FWGPU_LOGI << "WebGPU adapter " << toString(details)
FWGPU_LOGI << "WebGPU adapter " << details
<< " does not have the minimum required features.";
FWGPU_LOGI << "Missing required feature(s): ";
for (wgpu::FeatureName const& requiredFeature: REQUIRED_FEATURES) {
if (!details.adapter.HasFeature(requiredFeature)) {
FWGPU_LOGI << webGPUPrintableToString(requiredFeature);
FWGPU_LOGI << requiredFeature;
}
}
FWGPU_LOGI << "";
@@ -438,10 +458,10 @@ struct AdapterDetailsHash final {
}
wgpu::Limits supportedLimits {};
FILAMENT_CHECK_POSTCONDITION(details.adapter.GetLimits(&supportedLimits))
<< "Failed to get limits for WebGPU adapter: " << toString(details);
<< "Failed to get limits for WebGPU adapter: " << details;
if (!satisfiesLimits(supportedLimits)) {
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
FWGPU_LOGI << " (for WebGPU adapter " << toString(details) << ")";
FWGPU_LOGI << " (for WebGPU adapter " << details << ")";
#endif
return false;
}
@@ -467,15 +487,14 @@ struct AdapterDetailsHash final {
status != wgpu::RequestAdapterStatus::CallbackCancelled)
<< "Failed to request a WebGPU adapter due to the request callback "
"being cancelled? Options: "
<< adapterOptionsToString(options) << " " << message.data;
<< options << " " << message.data;
FILAMENT_CHECK_POSTCONDITION(status != wgpu::RequestAdapterStatus::Error)
<< "Failed to request a WebGPU adapter due to an error. Options: "
<< adapterOptionsToString(options) << " Error: " << message.data;
<< options << " Error: " << message.data;
if (status == wgpu::RequestAdapterStatus::Success) {
AdapterDetails details = AdapterDetails(readyAdapter);
FILAMENT_CHECK_POSTCONDITION(readyAdapter.GetInfo(&details.info))
<< "Failed to get info for adapter (options: "
<< adapterOptionsToString(options) << ")";
<< "Failed to get info for adapter (options: " << options << ")";
const std::lock_guard<std::mutex> lock(adaptersMutex);
compatibleAdapters.emplace(std::move(details.info), details.powerPreference,
std::move(details.adapter));
@@ -490,11 +509,9 @@ struct AdapterDetailsHash final {
wgpu::Future& future = futures[i];
wgpu::WaitStatus status = instance.WaitAny(future, REQUEST_ADAPTER_TIMEOUT_NANOSECONDS);
FILAMENT_CHECK_POSTCONDITION(status != wgpu::WaitStatus::TimedOut)
<< "Timed out requesting a WebGPU adapter with options "
<< adapterOptionsToString(options);
<< "Timed out requesting a WebGPU adapter with options " << options;
FILAMENT_CHECK_POSTCONDITION(status != wgpu::WaitStatus::Error)
<< "Failed to request a WebGPU adapter with options "
<< adapterOptionsToString(options)
<< "Failed to request a WebGPU adapter with options " << options
<< " due to an error (as request was made synchronous)";
assert_invariant(status == wgpu::WaitStatus::Success);
}
@@ -502,7 +519,7 @@ struct AdapterDetailsHash final {
#if FWGPU_ENABLED(FWGPU_PRINT_SYSTEM)
FWGPU_LOGI << compatibleAdapters.size() << " WebGPU adapter(s) found:";
for (auto& details: compatibleAdapters) {
FWGPU_LOGI << " WebGPU adapter: " << toString(details);
FWGPU_LOGI << " WebGPU adapter: " << details;
}
#endif
return compatibleAdapters;
@@ -569,7 +586,7 @@ void printDeviceDetails(wgpu::Device const& device) {
std::for_each(supportedFeatures.features,
supportedFeatures.features + supportedFeatures.featureCount,
[](wgpu::FeatureName const featureName) {
FWGPU_LOGI << " " << webGPUPrintableToString(featureName);
FWGPU_LOGI << " " << featureName;
});
}
wgpu::Limits supportedLimits{};
@@ -611,21 +628,7 @@ wgpu::Device WebGPUPlatform::requestDevice(wgpu::Adapter const& adapter) {
deviceDescriptor.defaultQueue.label = "default_queue";
deviceDescriptor.requiredFeatureCount = enabledFeatures.size();
deviceDescriptor.requiredFeatures = enabledFeatures.data();
// It is helpful to increase maxStorageTexturesPerShaderStage if the hardware supports it, but
// not required. This allows our mipmap generation to perform as well as possible for the given
// hardware. If we have more cases of "Nice to increase" limits like this we can generalize it
// in the future.
wgpu::Limits supportedLimits{};
FILAMENT_CHECK_POSTCONDITION(
adapter.GetLimits(&supportedLimits).status == wgpu::Status::Success)
<< "Failed to get limits for WebGPU adapter";
auto limitsToRequest = REQUIRED_LIMITS;
limitsToRequest.maxStorageTexturesPerShaderStage =
std::min(MAX_MIPMAP_STORAGE_TEXTURES_PER_STAGE,
supportedLimits.maxStorageTexturesPerShaderStage);
deviceDescriptor.requiredLimits = &limitsToRequest;
deviceDescriptor.requiredLimits = &REQUIRED_LIMITS;
deviceDescriptor.SetDeviceLostCallback(wgpu::CallbackMode::AllowSpontaneous,
[](wgpu::Device const&, wgpu::DeviceLostReason const& reason,
wgpu::StringView message) {
@@ -681,6 +684,7 @@ wgpu::Device WebGPUPlatform::requestDevice(wgpu::Adapter const& adapter) {
"adapter should support them: %s\n",
missingFeatures, featureNamesStream.str().data());
}
wgpu::Limits supportedLimits {};
FILAMENT_CHECK_POSTCONDITION(device.GetLimits(&supportedLimits))
<< "Failed to get limits for the device?";
FILAMENT_CHECK_POSTCONDITION(satisfiesLimits(supportedLimits))
+3 -15
View File
@@ -40,24 +40,18 @@ using namespace filament::math;
using namespace image;
#endif
#include <iostream>
namespace test {
Backend BackendTest::sBackend = Backend::NOOP;
OperatingSystem BackendTest::sOperatingSystem = OperatingSystem::OTHER;
bool BackendTest::sIsMobilePlatform = false;
int BackendTest::sArgc = 0;
char** BackendTest::sArgv = nullptr;
std::vector<std::string> BackendTest::sFailedImages;
void BackendTest::init(Backend backend, OperatingSystem operatingSystem, bool isMobilePlatform,
int argc, char** argv) {
void BackendTest::init(Backend backend, OperatingSystem operatingSystem, bool isMobilePlatform) {
sBackend = backend;
sOperatingSystem = operatingSystem;
sIsMobilePlatform = isMobilePlatform;
sArgc = argc;
sArgv = argv;
}
BackendTest::BackendTest() : commandBufferQueue(CONFIG_MIN_COMMAND_BUFFERS_SIZE,
@@ -166,11 +160,6 @@ void BackendTest::markImageAsFailure(std::string failedImageName) {
sFailedImages.emplace_back(std::move(failedImageName));
}
std::filesystem::path BackendTest::binaryDirectory() {
assert(sArgc >= 1);
return std::filesystem::path(sArgv[0]).remove_filename().string();
}
void BackendTest::recordFailedImages() {
if (!sFailedImages.empty()) {
std::string failedImages;
@@ -198,9 +187,8 @@ public:
}
};
void initTests(Backend backend, OperatingSystem operatingSystem, bool isMobile, int& argc,
char* argv[]) {
BackendTest::init(backend, operatingSystem, isMobile, argc, argv);
void initTests(Backend backend, OperatingSystem operatingSystem, bool isMobile, int& argc, char* argv[]) {
BackendTest::init(backend, operatingSystem, isMobile);
::testing::InitGoogleTest(&argc, argv);
::testing::AddGlobalTestEnvironment(new Environment);
}
+2 -8
View File
@@ -19,8 +19,6 @@
#include <gtest/gtest.h>
#include <filesystem>
#include <backend/Platform.h>
#include "private/backend/CommandBufferQueue.h"
@@ -33,20 +31,16 @@ namespace test {
class BackendTest : public ::testing::Test {
public:
static void init(Backend backend, OperatingSystem operatingSystem, bool isMobilePlatform,
int argc, char** argv);
static void init(Backend backend, OperatingSystem operatingSystem, bool isMobilePlatform);
static Backend sBackend;
static OperatingSystem sOperatingSystem;
static bool sIsMobilePlatform;
static int sArgc;
static char** sArgv;
// Takes the name of the image that wasn't correct, without the .png suffix
static void markImageAsFailure(std::string failedImageName);
static std::filesystem::path binaryDirectory();
protected:
BackendTest();
+8 -8
View File
@@ -59,28 +59,28 @@ uint32_t ScreenshotParams::expectedHash() const {
return mExpectedPixelHash;
}
std::filesystem::path ScreenshotParams::actualDirectoryPath() {
return BackendTest::binaryDirectory().append("images/actual_images");
std::string ScreenshotParams::actualDirectoryPath() {
return "images/actual_images";
}
std::string ScreenshotParams::actualFileName() const {
return absl::StrFormat("%s_actual.png", mFileName);
}
std::filesystem::path ScreenshotParams::actualFilePath() const {
return actualDirectoryPath().append(actualFileName());
std::string ScreenshotParams::actualFilePath() const {
return absl::StrFormat("%s/%s", actualDirectoryPath(), actualFileName());
}
std::filesystem::path ScreenshotParams::expectedDirectoryPath() {
return BackendTest::binaryDirectory().append("images/expected_images");
std::string ScreenshotParams::expectedDirectoryPath() {
return "images/expected_images";
}
std::string ScreenshotParams::expectedFileName() const {
return absl::StrFormat("%s.png", mFileName);
}
std::filesystem::path ScreenshotParams::expectedFilePath() const {
return expectedDirectoryPath().append(expectedFileName());
std::string ScreenshotParams::expectedFilePath() const {
return absl::StrFormat("%s/%s", expectedDirectoryPath(), expectedFileName());
}
const std::string ScreenshotParams::filePrefix() const {
+4 -5
View File
@@ -17,7 +17,6 @@
#ifndef TNT_IMAGE_EXPECTATIONS_H
#define TNT_IMAGE_EXPECTATIONS_H
#include <filesystem>
#include <vector>
#include "gtest/gtest.h"
@@ -52,12 +51,12 @@ public:
bool isSrgb() const;
uint32_t expectedHash() const;
static std::filesystem::path actualDirectoryPath();
static std::string actualDirectoryPath();
std::string actualFileName() const;
std::filesystem::path actualFilePath() const;
static std::filesystem::path expectedDirectoryPath();
std::string actualFilePath() const;
static std::string expectedDirectoryPath();
std::string expectedFileName() const;
std::filesystem::path expectedFilePath() const;
std::string expectedFilePath() const;
const std::string filePrefix() const;
private:
-20
View File
@@ -139,24 +139,6 @@ TEST_F(ReadPixelsTest, ReadPixels) {
return bufferDimension;
}
void exportScreenshot(void* pixelData) const {
#ifndef FILAMENT_IOS
const size_t width = readRect.width, height = readRect.height;
LinearImage image(width, height, 4);
if (format == PixelDataFormat::RGBA && type == PixelDataType::UBYTE) {
image = toLinearWithAlpha<uint8_t>(width, height, width * 4, (uint8_t*)pixelData);
}
if (format == PixelDataFormat::RGBA && type == PixelDataType::FLOAT) {
memcpy(image.getPixelRef(), pixelData, width * height * sizeof(math::float4));
}
std::string png = std::string(testName) + ".png";
std::filesystem::path path = ScreenshotParams::actualDirectoryPath();
path.append(png);
std::ofstream outputStream(path.c_str(), std::ios::binary | std::ios::trunc);
ImageEncoder::encode(outputStream, ImageEncoder::Format::PNG, image, "", png);
#endif
}
// The format and type for the readPixels call.
PixelDataFormat format = PixelDataFormat::RGBA;
PixelDataType type = PixelDataType::UBYTE;
@@ -329,8 +311,6 @@ TEST_F(ReadPixelsTest, ReadPixels) {
const auto* test = (const TestCase*)user;
assert_invariant(test);
test->exportScreenshot(buffer);
// Hash the contents of the buffer and check that they match.
uint32_t hash = utils::hash::murmur3((const uint32_t*)buffer, size / 4, 0);
-9
View File
@@ -317,15 +317,6 @@ public:
*/
size_t metalUploadBufferSizeBytes = 512 * 1024;
/**
* The action to take if a Drawable cannot be acquired.
*
* Each frame rendered requires a CAMetalDrawable texture, which is
* presented on-screen at the completion of each frame. These are
* limited and provided round-robin style by the system.
*/
bool metalDisablePanicOnDrawableFailure = false;
/**
* Set to `true` to forcibly disable parallel shader compilation in the backend.
* Currently only honored by the GL and Metal backends.
+53 -3
View File
@@ -86,6 +86,9 @@ public:
std::is_same_v<math::mat3f, T>
>;
template<typename T>
using is_supported_constant_parameter_t = std::enable_if_t<std::is_same_v<bool, T>>;
/**
* Creates a new MaterialInstance using another MaterialInstance as a template for initialization.
* The new MaterialInstance is an instance of the same Material of the template instance and
@@ -237,13 +240,13 @@ public:
/**
* Gets the value of a parameter by name.
*
*
* Note: Only supports non-texture parameters such as numeric and math types.
*
*
* @param name Name of the parameter as defined by Material. Cannot be nullptr.
* @param nameLength Length in `char` of the name parameter.
* @throws utils::PreConditionPanic if name doesn't exist or no-op if exceptions are disabled.
*
*
* @see Material::hasParameter
*/
template<typename T>
@@ -261,6 +264,53 @@ public:
return getParameter<T>(name, strlen(name));
}
/**
* Set a mutable constant parameter by name.
*
* @param name Name of the parameter as defined by Material. Cannot be nullptr.
* @param nameLength Length in `char` of the name parameter.
* @param value Value of the parameter to set.
* @throws utils::PreConditionPanic if name doesn't exist or no-op if exceptions are disabled.
*/
template<typename T, typename = is_supported_constant_parameter_t<T>>
void setConstant(const char* UTILS_NONNULL name, size_t nameLength, T const& value);
/** inline helper to provide the name as a null-terminated string literal */
template<typename T, typename = is_supported_constant_parameter_t<T>>
void setConstant(StringLiteral const name, T const& value) {
setConstant<T>(name.data, name.size, value);
}
/** inline helper to provide the name as a null-terminated C string */
template<typename T, typename = is_supported_constant_parameter_t<T>>
void setConstant(const char* UTILS_NONNULL name, T const& value) {
setConstant<T>(name, strlen(name), value);
}
/**
* Gets the value of a mutable constant parameter by name.
*
* @param name Name of the parameter as defined by Material. Cannot be nullptr.
* @param nameLength Length in `char` of the name parameter.
* @throws utils::PreConditionPanic if name doesn't exist or no-op if exceptions are disabled.
*
* @see Material::hasConstant
*/
template<typename T, typename = is_supported_constant_parameter_t<T>>
T getConstant(const char* UTILS_NONNULL name, size_t nameLength) const;
/** inline helper to provide the name as a null-terminated C string */
template<typename T, typename = is_supported_constant_parameter_t<T>>
T getConstant(StringLiteral const name) const {
return getConstant<T>(name.data, name.size);
}
/** inline helper to provide the name as a null-terminated C string */
template<typename T, typename = is_supported_constant_parameter_t<T>>
T getConstant(const char* UTILS_NONNULL name) const {
return getConstant<T>(name, strlen(name));
}
/**
* Set-up a custom scissor rectangle; by default it is disabled.
*
+40
View File
@@ -232,6 +232,46 @@ template UTILS_PUBLIC mat3f MaterialInstance::getParameter<mat3f> (const ch
// ------------------------------------------------------------------------------------------------
template<>
inline void FMaterialInstance::setConstantImpl<bool>(std::string_view const name, bool const& value) {
std::optional<uint32_t> id = mMaterial->getMutableConstantId(name);
FILAMENT_CHECK_PRECONDITION(id.has_value()) << "No mutable constant with name " << name;
mConstants.set(*id, value);
}
template<typename T, typename>
void MaterialInstance::setConstant(const char* name, size_t nameLength, T const& value) {
downcast(this)->setConstantImpl({ name, nameLength }, value);
}
// Explicit template instantiation of our supported types.
//
// Mutable spec constants will probably only ever allow bools, but it's nice to keep the API
// forwards-compatible.
template UTILS_PUBLIC void MaterialInstance::setConstant<bool>(const char* name, size_t nameLength, bool const& v);
// ------------------------------------------------------------------------------------------------
template<>
inline bool FMaterialInstance::getConstantImpl<bool>(std::string_view const name) const {
std::optional<uint32_t> id = mMaterial->getMutableConstantId(name);
FILAMENT_CHECK_PRECONDITION(id.has_value()) << "No mutable constant with name " << name;
return mConstants[*id];
}
template<typename T, typename>
T MaterialInstance::getConstant(const char* name, size_t nameLength) const {
return downcast(this)->getConstantImpl<T>({ name, nameLength });
}
// Explicit template instantiation of our supported types.
//
// Mutable spec constants will probably only ever allow bools, but it's nice to keep the API
// forwards-compatible.
template UTILS_PUBLIC bool MaterialInstance::getConstant<bool>(const char* name, size_t nameLength) const;
// ------------------------------------------------------------------------------------------------
Material const* MaterialInstance::getMaterial() const noexcept {
return downcast(this)->getMaterial();
}
+37 -6
View File
@@ -229,6 +229,11 @@ bool MaterialParser::getConstants(FixedCapacityVector<MaterialConstant>* contain
return get<ChunkMaterialConstants>(container);
}
bool MaterialParser::getMutableConstants(
FixedCapacityVector<MaterialMutableConstant>* container) const noexcept {
return get<ChunkMaterialMutableConstants>(container);
}
bool MaterialParser::getPushConstants(CString* structVarName,
FixedCapacityVector<MaterialPushConstant>* value) const noexcept {
auto [start, end] = mImpl.mChunkContainer.getChunkRange(MaterialPushConstants);
@@ -481,7 +486,6 @@ bool ChunkSamplerInterfaceBlock::unflatten(Unflattener& unflattener,
uint8_t fieldType = 0;
uint8_t fieldFormat = 0;
uint8_t fieldPrecision = 0;
bool fieldUnfilterable = false;
bool fieldMultisample = false;
if (!unflattener.read(&fieldName)) {
@@ -504,10 +508,6 @@ bool ChunkSamplerInterfaceBlock::unflatten(Unflattener& unflattener,
return false;
}
if (!unflattener.read(&fieldUnfilterable)) {
return false;
}
if (!unflattener.read(&fieldMultisample)) {
return false;
}
@@ -517,7 +517,6 @@ bool ChunkSamplerInterfaceBlock::unflatten(Unflattener& unflattener,
SamplerInterfaceBlock::Type(fieldType),
SamplerInterfaceBlock::Format(fieldFormat),
SamplerInterfaceBlock::Precision(fieldPrecision),
fieldUnfilterable,
fieldMultisample);
}
@@ -757,6 +756,38 @@ bool ChunkMaterialConstants::unflatten(Unflattener& unflattener,
return true;
}
bool ChunkMaterialMutableConstants::unflatten(Unflattener& unflattener,
FixedCapacityVector<MaterialMutableConstant>* materialConstants) {
assert_invariant(materialConstants);
// Read number of constants.
uint64_t numConstants = 0;
if (!unflattener.read(&numConstants)) {
return false;
}
materialConstants->reserve(numConstants);
materialConstants->resize(numConstants);
for (uint64_t i = 0; i < numConstants; i++) {
CString constantName;
bool defaultValue;
if (!unflattener.read(&constantName)) {
return false;
}
if (!unflattener.read(&defaultValue)) {
return false;
}
(*materialConstants)[i].name = std::move(constantName);
(*materialConstants)[i].defaultValue = defaultValue;
}
return true;
}
bool ChunkMaterialPushConstants::unflatten(Unflattener& unflattener,
CString* structVarName,
FixedCapacityVector<MaterialPushConstant>* materialPushConstants) {
+10
View File
@@ -49,6 +49,7 @@ class BufferInterfaceBlock;
class SamplerInterfaceBlock;
struct SubpassInfo;
struct MaterialConstant;
struct MaterialMutableConstant;
struct MaterialPushConstant;
class MaterialParser {
@@ -79,6 +80,8 @@ public:
bool getShaderModels(uint32_t* value) const noexcept;
bool getMaterialProperties(uint64_t* value) const noexcept;
bool getConstants(utils::FixedCapacityVector<MaterialConstant>* value) const noexcept;
bool getMutableConstants(utils::FixedCapacityVector<MaterialMutableConstant>* value)
const noexcept;
bool getPushConstants(utils::CString* structVarName,
utils::FixedCapacityVector<MaterialPushConstant>* value) const noexcept;
@@ -237,6 +240,13 @@ struct ChunkMaterialConstants {
static filamat::ChunkType const tag = filamat::MaterialConstants;
};
struct ChunkMaterialMutableConstants {
static bool unflatten(filaflat::Unflattener& unflattener,
utils::FixedCapacityVector<MaterialMutableConstant>* materialConstants);
using Container = utils::FixedCapacityVector<MaterialMutableConstant>;
static filamat::ChunkType const tag = filamat::MaterialMutableConstants;
};
struct ChunkMaterialPushConstants {
static bool unflatten(filaflat::Unflattener& unflattener, utils::CString* structVarName,
utils::FixedCapacityVector<MaterialPushConstant>* materialPushConstants);
+8 -2
View File
@@ -187,7 +187,10 @@ FMaterial* PostProcessManager::PostProcessMaterial::getMaterial(FEngine& engine,
if (UTILS_UNLIKELY(mSize)) {
loadMaterial(engine);
}
mMaterial->prepareProgram(Variant{ Variant::type_t(variant) });
mMaterial->prepareProgram(Variant{ Variant::type_t(variant) },
// TODO(exv): allow post process materials to use mutable spec constants?
Program::MutableSpecConstantsInfo(0)
);
return mMaterial;
}
@@ -423,7 +426,10 @@ UTILS_NOINLINE
PipelineState PostProcessManager::getPipelineState(
FMaterial const* const ma, PostProcessVariant variant) const noexcept {
return {
.program = ma->getProgram(Variant{ Variant::type_t(variant) }),
.program = ma->getProgram(Variant{ Variant::type_t(variant) },
// TODO(exv): allow post process materials to use mutable spec constants?
Program::MutableSpecConstantsInfo(0)
),
.vertexBufferInfo = mFullScreenQuadVbih,
.pipelineLayout = {
.setLayout = {
+4 -2
View File
@@ -246,7 +246,8 @@ void RenderPass::appendCommands(FEngine const& engine,
for (Command const* first = curr, *last = curr + commandCount ; first != last ; ++first) {
if (UTILS_LIKELY((first->key & CUSTOM_MASK) == uint64_t(CustomCommand::PASS))) {
auto ma = first->info.mi->getMaterial();
ma->prepareProgram(first->info.materialVariant);
ma->prepareProgram(first->info.materialVariant,
first->info.mi->getMutableSpecConstants());
}
}
}
@@ -1098,7 +1099,8 @@ void RenderPass::Executor::execute(FEngine const& engine, DriverApi& driver,
}
assert_invariant(ma);
pipeline.program = ma->getProgram(info.materialVariant);
pipeline.program = ma->getProgram(info.materialVariant,
info.mi->getMutableSpecConstants());
if (UTILS_UNLIKELY(memcmp(&pipeline, &currentPipeline, sizeof(PipelineState)) != 0)) {
currentPipeline = pipeline;
-2
View File
@@ -140,7 +140,6 @@ Engine* FEngine::create(Builder const& builder) {
.forceGLES2Context = instance->getConfig().forceGLES2Context,
.stereoscopicType = instance->getConfig().stereoscopicType,
.assertNativeWindowIsValid = instance->features.backend.opengl.assert_native_window_is_valid,
.metalDisablePanicOnDrawableFailure = instance->getConfig().metalDisablePanicOnDrawableFailure,
};
instance->mDriver = platform->createDriver(sharedContext, driverConfig);
@@ -760,7 +759,6 @@ int FEngine::loop() {
.forceGLES2Context = mConfig.forceGLES2Context,
.stereoscopicType = mConfig.stereoscopicType,
.assertNativeWindowIsValid = features.backend.opengl.assert_native_window_is_valid,
.metalDisablePanicOnDrawableFailure = mConfig.metalDisablePanicOnDrawableFailure,
};
mDriver = mPlatform->createDriver(mSharedGLContext, driverConfig);
+100 -32
View File
@@ -342,8 +342,24 @@ FMaterial::FMaterial(FEngine& engine, const Builder& builder,
processBlendingMode(parser);
processSpecializationConstants(engine, builder, parser);
processMutableSpecializationConstants(engine, builder, parser);
processPushConstants(engine, parser);
processDescriptorSets(engine, parser);
switch (mMaterialDomain) {
case filament::MaterialDomain::SURFACE:
mVariantBits = VARIANT_BITS;
break;
case filament::MaterialDomain::POST_PROCESS:
mVariantBits = POST_PROCESS_VARIANT_BITS;
break;
case filament::MaterialDomain::COMPUTE:
mVariantBits = 0;
break;
}
mCachedPrograms = FixedCapacityVector<backend::Handle<backend::HwProgram>>(
1 << (mVariantBits + mMaterialMutableConstants.size()));
precacheDepthVariants(engine);
#if FILAMENT_ENABLE_MATDBG
@@ -449,13 +465,17 @@ void FMaterial::compile(CompilerPriorityQueue const priority,
UserVariantFilterMask const variantFilter =
~variantSpec & UserVariantFilterMask(UserVariantFilterBit::ALL);
int mutableConstantVariations = 1 << mMaterialMutableConstants.size();
if (UTILS_LIKELY(mEngine.getDriverApi().isParallelShaderCompileSupported())) {
auto const& variants = isVariantLit() ?
VariantUtils::getLitVariants() : VariantUtils::getUnlitVariants();
for (auto const variant: variants) {
if (!variantFilter || variant == Variant::filterUserVariant(variant, variantFilter)) {
if (hasVariant(variant)) {
prepareProgram(variant, priority);
for (int spec = 0; spec < mutableConstantVariations; spec++) {
prepareProgram(variant, Program::MutableSpecConstantsInfo(spec), priority);
}
}
}
}
@@ -496,6 +516,11 @@ FMaterialInstance* FMaterial::getDefaultInstance() noexcept {
return mDefaultMaterialInstance;
}
bool FMaterial::isCached(Variant const variant,
const backend::Program::MutableSpecConstantsInfo mutableSpecConstants) const noexcept {
return bool(mCachedPrograms[getCachedProgramIndex(variant, mutableSpecConstants)]);
}
bool FMaterial::hasParameter(const char* name) const noexcept {
return mUniformInterfaceBlock.hasField(name) ||
mSamplerInterfaceBlock.hasSampler(name) ||
@@ -536,14 +561,15 @@ bool FMaterial::hasVariant(Variant const variant) const noexcept {
}
void FMaterial::prepareProgramSlow(Variant const variant,
const backend::Program::MutableSpecConstantsInfo mutableSpecConstants,
backend::CompilerPriorityQueue const priorityQueue) const noexcept {
assert_invariant(mEngine.hasFeatureLevel(mFeatureLevel));
switch (getMaterialDomain()) {
case MaterialDomain::SURFACE:
getSurfaceProgramSlow(variant, priorityQueue);
getSurfaceProgramSlow(variant, mutableSpecConstants, priorityQueue);
break;
case MaterialDomain::POST_PROCESS:
getPostProcessProgramSlow(variant, priorityQueue);
getPostProcessProgramSlow(variant, mutableSpecConstants, priorityQueue);
break;
case MaterialDomain::COMPUTE:
// TODO: implement MaterialDomain::COMPUTE
@@ -552,6 +578,7 @@ void FMaterial::prepareProgramSlow(Variant const variant,
}
void FMaterial::getSurfaceProgramSlow(Variant const variant,
const backend::Program::MutableSpecConstantsInfo mutableSpecConstants,
CompilerPriorityQueue const priorityQueue) const noexcept {
// filterVariant() has already been applied in generateCommands(), shouldn't be needed here
// if we're unlit, we don't have any bits that correspond to lit materials
@@ -562,25 +589,25 @@ void FMaterial::getSurfaceProgramSlow(Variant const variant,
Variant const vertexVariant = Variant::filterVariantVertex(variant);
Variant const fragmentVariant = Variant::filterVariantFragment(variant);
Program pb{ getProgramWithVariants(variant, vertexVariant, fragmentVariant) };
Program pb{ getProgramWithVariants(variant, vertexVariant, fragmentVariant, mutableSpecConstants) };
pb.priorityQueue(priorityQueue);
pb.multiview(
mEngine.getConfig().stereoscopicType == StereoscopicType::MULTIVIEW &&
Variant::isStereoVariant(variant));
createAndCacheProgram(std::move(pb), variant);
createAndCacheProgram(std::move(pb), variant, mutableSpecConstants);
}
void FMaterial::getPostProcessProgramSlow(Variant const variant,
const backend::Program::MutableSpecConstantsInfo mutableSpecConstants,
CompilerPriorityQueue const priorityQueue) const noexcept {
Program pb{ getProgramWithVariants(variant, variant, variant) };
Program pb{ getProgramWithVariants(variant, variant, variant, mutableSpecConstants) };
pb.priorityQueue(priorityQueue);
createAndCacheProgram(std::move(pb), variant);
createAndCacheProgram(std::move(pb), variant, mutableSpecConstants);
}
Program FMaterial::getProgramWithVariants(
Variant variant,
Variant vertexVariant,
Variant fragmentVariant) const {
Program FMaterial::getProgramWithVariants(Variant variant,
Variant vertexVariant, Variant fragmentVariant,
const backend::Program::MutableSpecConstantsInfo mutableSpecConstants) const {
FEngine const& engine = mEngine;
const ShaderModel sm = engine.getShaderModel();
const bool isNoop = engine.getBackend() == Backend::NOOP;
@@ -641,29 +668,34 @@ Program FMaterial::getProgramWithVariants(
mProgramDescriptorBindings[+DescriptorSetBindingPoints::PER_RENDERABLE]);
program.descriptorBindings(+DescriptorSetBindingPoints::PER_MATERIAL,
mProgramDescriptorBindings[+DescriptorSetBindingPoints::PER_MATERIAL]);
program.specializationConstants(mSpecializationConstants);
program.specializationConstants(mSpecializationConstants,
CONFIG_MAX_RESERVED_SPEC_CONSTANTS + mMaterialConstants.size(), mutableSpecConstants);
program.pushConstants(ShaderStage::VERTEX, mPushConstants[uint8_t(ShaderStage::VERTEX)]);
program.pushConstants(ShaderStage::FRAGMENT, mPushConstants[uint8_t(ShaderStage::FRAGMENT)]);
program.cacheId(hash::combine(size_t(mCacheId), variant.key));
program.cacheId(
hash::combine(size_t(mCacheId),
hash::combine(variant.key, mutableSpecConstants.getValue())));
return program;
}
void FMaterial::createAndCacheProgram(Program&& p, Variant const variant) const noexcept {
void FMaterial::createAndCacheProgram(Program&& p, Variant const variant,
Program::MutableSpecConstantsInfo const mutableSpecConstants) const noexcept {
FEngine const& engine = mEngine;
DriverApi& driverApi = mEngine.getDriverApi();
bool const isShared = isSharedVariant(variant);
size_t const cacheIndex = getCachedProgramIndex(variant, mutableSpecConstants);
// Check if the default material has this program cached
if (isShared) {
FMaterial const* const pDefaultMaterial = engine.getDefaultMaterial();
if (pDefaultMaterial) {
auto const program = pDefaultMaterial->mCachedPrograms[variant.key];
auto const program = pDefaultMaterial->mCachedPrograms[cacheIndex];
if (program) {
mCachedPrograms[variant.key] = program;
mCachedPrograms[cacheIndex] = program;
return;
}
}
@@ -672,17 +704,17 @@ void FMaterial::createAndCacheProgram(Program&& p, Variant const variant) const
auto const program = driverApi.createProgram(std::move(p));
driverApi.setDebugTag(program.getId(), mName);
assert_invariant(program);
mCachedPrograms[variant.key] = program;
mCachedPrograms[cacheIndex] = program;
// If the default material doesn't already have this program cached, and all caching conditions
// are met (Surface Domain and no custom depth shader), cache it now.
// New Materials will inherit these program automatically.
if (isShared) {
FMaterial const* const pDefaultMaterial = engine.getDefaultMaterial();
if (pDefaultMaterial && !pDefaultMaterial->mCachedPrograms[variant.key]) {
if (pDefaultMaterial && !pDefaultMaterial->mCachedPrograms[cacheIndex]) {
// set the tag to the default material name
driverApi.setDebugTag(program.getId(), mName);
pDefaultMaterial->mCachedPrograms[variant.key] = program;
pDefaultMaterial->mCachedPrograms[cacheIndex] = program;
}
}
}
@@ -730,6 +762,14 @@ size_t FMaterial::getParameters(ParameterInfo* parameters, size_t count) const n
return count;
}
std::optional<uint32_t> FMaterial::getMutableConstantId(std::string_view name) const noexcept {
auto const pos = mMutableSpecializationConstantsNameToIndex.find(name);
if (pos != mMutableSpecializationConstantsNameToIndex.end()) {
return pos->second;
}
return std::nullopt;
}
#if FILAMENT_ENABLE_MATDBG
// Swaps in an edited version of the original package that was used to create the material. The
@@ -788,7 +828,9 @@ void FMaterial::onQueryCallback(void* userdata, VariantList* pActiveVariants) {
#endif // FILAMENT_ENABLE_MATDBG
[[nodiscard]] Handle<HwProgram> FMaterial::getProgramWithMATDBG(Variant const variant) const noexcept {
[[nodiscard]] Handle<HwProgram> FMaterial::getProgramWithMATDBG(Variant const variant,
Program::MutableSpecConstantsInfo const mutableSpecConstants) const noexcept {
size_t cacheIndex = getCachedProgramIndex(variant, mutableSpecConstants);
#if FILAMENT_ENABLE_MATDBG
assert_invariant((size_t)variant.key < VARIANT_COUNT);
std::unique_lock lock(mActiveProgramsLock);
@@ -803,13 +845,13 @@ void FMaterial::onQueryCallback(void* userdata, VariantList* pActiveVariants) {
lock.unlock();
if (isSharedVariant(variant)) {
FMaterial const* const pDefaultMaterial = mEngine.getDefaultMaterial();
if (pDefaultMaterial && pDefaultMaterial->mCachedPrograms[variant.key]) {
return pDefaultMaterial->getProgram(variant);
if (pDefaultMaterial && pDefaultMaterial->mCachedPrograms[cacheIndex]) {
return pDefaultMaterial->getProgram(variant, mutableSpecConstants);
}
}
#endif
assert_invariant(mCachedPrograms[variant.key]);
return mCachedPrograms[variant.key];
assert_invariant(mCachedPrograms[cacheIndex]);
return mCachedPrograms[cacheIndex];
}
void FMaterial::destroyPrograms(FEngine& engine,
@@ -822,7 +864,7 @@ void FMaterial::destroyPrograms(FEngine& engine,
case MaterialDomain::SURFACE: {
if (mIsDefaultMaterial || mHasCustomDepthShader) {
// default material, or we have custom depth shaders, we destroy all variants
for (size_t k = 0, n = VARIANT_COUNT; k < n; ++k) {
for (size_t k = 0, n = cachedPrograms.size(); k < n; ++k) {
if ((k & variantMask) == variantValue) {
// Only destroy if the handle is valid. Not strictly needed, but we have a lot
// of variants, and this generates traffic in the command queue.
@@ -840,7 +882,7 @@ void FMaterial::destroyPrograms(FEngine& engine,
UTILS_UNUSED_IN_RELEASE
auto const UTILS_NULLABLE pDefaultMaterial = engine.getDefaultMaterial();
for (size_t k = 0, n = VARIANT_COUNT; k < n; ++k) {
for (size_t k = 0, n = cachedPrograms.size(); k < n; ++k) {
if ((k & variantMask) == variantValue) {
// Only destroy if the handle is valid. Not strictly needed, but we have a lot
// of variant, and this generates traffic in the command queue.
@@ -852,7 +894,7 @@ void FMaterial::destroyPrograms(FEngine& engine,
// During Engine::shutdown, auto-cleanup destroys the
// default material first
assert_invariant(!pDefaultMaterial ||
pDefaultMaterial->mCachedPrograms[k]);
pDefaultMaterial->mCachedPrograms[k & (VARIANT_COUNT - 1)]);
// we don't own this variant, skip, but clear the entry.
cachedPrograms[k].clear();
continue;
@@ -866,7 +908,7 @@ void FMaterial::destroyPrograms(FEngine& engine,
break;
}
case MaterialDomain::POST_PROCESS: {
for (size_t k = 0, n = POST_PROCESS_VARIANT_COUNT; k < n; ++k) {
for (size_t k = 0, n = cachedPrograms.size(); k < n; ++k) {
if ((k & variantMask) == variantValue) {
// Only destroy if the handle is valid. Not strictly needed, but we have a lot
// of variant, and this generates traffic in the command queue.
@@ -878,8 +920,13 @@ void FMaterial::destroyPrograms(FEngine& engine,
break;
}
case MaterialDomain::COMPUTE: {
// Compute programs don't have variants
driverApi.destroyProgram(std::move(cachedPrograms[0]));
for (size_t k = 0, n = cachedPrograms.size(); k < n; ++k) {
// Only destroy if the handle is valid. Not strictly needed, but we have a lot
// of variant, and this generates traffic in the command queue.
if (cachedPrograms[k]) {
driverApi.destroyProgram(std::move(cachedPrograms[k]));
}
}
break;
}
}
@@ -1114,6 +1161,19 @@ void FMaterial::processSpecializationConstants(FEngine& engine, Builder const& b
}
}
void FMaterial::processMutableSpecializationConstants(FEngine& engine, Builder const& builder,
MaterialParser const* const parser) {
// Older materials won't have a mutable constants chunk, but that's okay.
parser->getMutableConstants(&mMaterialMutableConstants);
for (size_t i = 0, c = mMaterialMutableConstants.size(); i < c; i++) {
auto& item = mMaterialMutableConstants[i];
// the key can be a string_view because mMaterialMutableConstants owns the CString
std::string_view const key{ item.name.data(), item.name.size() };
mMutableSpecializationConstantsNameToIndex[key] = i;
mDefaultMutableSpecializationConstants.set(i, item.defaultValue);
}
}
void FMaterial::processPushConstants(FEngine&, MaterialParser const* parser) {
FixedCapacityVector<Program::PushConstant>& vertexConstants =
mPushConstants[uint8_t(ShaderStage::VERTEX)];
@@ -1169,7 +1229,8 @@ void FMaterial::precacheDepthVariants(FEngine& engine) {
}
assert_invariant(Variant::isValidDepthVariant(variant));
if (hasVariant(variant)) {
prepareProgram(variant);
// The default material shouldn't ever have mutable spec constants.
prepareProgram(variant, Program::MutableSpecConstantsInfo(0));
}
}
return;
@@ -1182,9 +1243,16 @@ void FMaterial::precacheDepthVariants(FEngine& engine) {
FMaterial const* const pDefaultMaterial = engine.getDefaultMaterial();
assert_invariant(pDefaultMaterial);
auto const allDepthVariants = VariantUtils::getDepthVariants();
int mutableConstantVariations = 1 << mMaterialMutableConstants.size();
for (auto const variant: allDepthVariants) {
assert_invariant(Variant::isValidDepthVariant(variant));
mCachedPrograms[variant.key] = pDefaultMaterial->mCachedPrograms[variant.key];
size_t defaultCacheIndex = pDefaultMaterial->getCachedProgramIndex(variant,
Program::MutableSpecConstantsInfo(0));
for (int spec = 0; spec < mutableConstantVariations; spec++) {
size_t cacheIndex = getCachedProgramIndex(variant,
Program::MutableSpecConstantsInfo(spec));
mCachedPrograms[cacheIndex] = pDefaultMaterial->mCachedPrograms[defaultCacheIndex];
}
}
}
}
+46 -14
View File
@@ -122,9 +122,8 @@ public:
FEngine& getEngine() const noexcept { return mEngine; }
bool isCached(Variant const variant) const noexcept {
return bool(mCachedPrograms[variant.key]);
}
bool isCached(Variant const variant,
const backend::Program::MutableSpecConstantsInfo mutableSpecConstants) const noexcept;
void invalidate(Variant::type_t variantMask = 0, Variant::type_t variantValue = 0) noexcept;
@@ -132,26 +131,31 @@ public:
// Must be called outside of backend render pass.
// Must be called before getProgram() below.
void prepareProgram(Variant const variant,
backend::Program::MutableSpecConstantsInfo const mutableSpecConstants,
backend::CompilerPriorityQueue const priorityQueue = CompilerPriorityQueue::HIGH) const noexcept {
// prepareProgram() is called for each RenderPrimitive in the scene, so it must be efficient.
if (UTILS_UNLIKELY(!isCached(variant))) {
prepareProgramSlow(variant, priorityQueue);
if (UTILS_UNLIKELY(!isCached(variant, mutableSpecConstants))) {
prepareProgramSlow(variant, mutableSpecConstants, priorityQueue);
}
}
// getProgram returns the backend program for the material's given variant.
// getProgram returns the backend program for the material's given variant and set of mutable
// spec constants.
// Must be called after prepareProgram().
[[nodiscard]]
backend::Handle<backend::HwProgram> getProgram(Variant const variant) const noexcept {
backend::Handle<backend::HwProgram> getProgram(Variant const variant,
backend::Program::MutableSpecConstantsInfo const mutableSpecConstants) const noexcept {
#if FILAMENT_ENABLE_MATDBG
return getProgramWithMATDBG(variant);
return getProgramWithMATDBG(variant, mutableSpecConstants);
#endif
assert_invariant(mCachedPrograms[variant.key]);
return mCachedPrograms[variant.key];
size_t index = getCachedProgramIndex(variant, mutableSpecConstants);
assert_invariant(mCachedPrograms[index]);
return mCachedPrograms[index];
}
[[nodiscard]]
backend::Handle<backend::HwProgram> getProgramWithMATDBG(Variant variant) const noexcept;
backend::Handle<backend::HwProgram> getProgramWithMATDBG(Variant variant,
backend::Program::MutableSpecConstantsInfo const mutableSpecConstants) const noexcept;
bool isVariantLit() const noexcept { return mIsVariantLit; }
@@ -208,6 +212,12 @@ public:
}
size_t getParameters(ParameterInfo* parameters, size_t count) const noexcept;
uint32_t getMutableConstantCount() const noexcept { return mMaterialMutableConstants.size(); }
std::optional<uint32_t> getMutableConstantId(std::string_view name) const noexcept;
backend::Program::MutableSpecConstantsInfo getDefaultMutableConstants() const noexcept {
return mDefaultMutableSpecializationConstants;
}
uint32_t generateMaterialInstanceId() const noexcept { return mMaterialInstanceId++; }
void destroyPrograms(FEngine& engine,
@@ -261,34 +271,48 @@ public:
private:
bool hasVariant(Variant variant) const noexcept;
void prepareProgramSlow(Variant variant,
backend::Program::MutableSpecConstantsInfo mutableSpecConstants,
CompilerPriorityQueue priorityQueue) const noexcept;
void getSurfaceProgramSlow(Variant variant,
backend::Program::MutableSpecConstantsInfo mutableSpecConstants,
CompilerPriorityQueue priorityQueue) const noexcept;
void getPostProcessProgramSlow(Variant variant,
backend::Program::MutableSpecConstantsInfo mutableSpecConstants,
CompilerPriorityQueue priorityQueue) const noexcept;
backend::Program getProgramWithVariants(Variant variant,
Variant vertexVariant, Variant fragmentVariant) const;
Variant vertexVariant, Variant fragmentVariant,
backend::Program::MutableSpecConstantsInfo mutableSpecConstants) const;
void processBlendingMode(MaterialParser const* parser);
void processSpecializationConstants(FEngine& engine, Builder const& builder,
MaterialParser const* parser);
void processMutableSpecializationConstants(FEngine& engine, Builder const& builder,
MaterialParser const* parser);
void processPushConstants(FEngine& engine, MaterialParser const* parser);
void precacheDepthVariants(FEngine& engine);
void processDescriptorSets(FEngine& engine, MaterialParser const* parser);
void createAndCacheProgram(backend::Program&& p, Variant variant) const noexcept;
void createAndCacheProgram(backend::Program&& p, Variant variant,
backend::Program::MutableSpecConstantsInfo mutableSpecConstants) const noexcept;
inline bool isSharedVariant(Variant const variant) const {
return (mMaterialDomain == MaterialDomain::SURFACE) && !mIsDefaultMaterial &&
!mHasCustomDepthShader && Variant::isValidDepthVariant(variant);
}
inline size_t getCachedProgramIndex(const Variant variant,
const backend::Program::MutableSpecConstantsInfo mutableSpecConstants) const noexcept {
return static_cast<size_t>(variant.key) |
(static_cast<size_t>(mutableSpecConstants.getValue()) << mVariantBits);
}
// try to order by frequency of use
mutable std::array<backend::Handle<backend::HwProgram>, VARIANT_COUNT> mCachedPrograms;
mutable utils::FixedCapacityVector<backend::Handle<backend::HwProgram>> mCachedPrograms;
DescriptorSetLayout mPerViewDescriptorSetLayout;
DescriptorSetLayout mPerViewDescriptorSetLayoutVsm;
DescriptorSetLayout mDescriptorSetLayout;
@@ -324,6 +348,7 @@ private:
bool mHasCustomDepthShader = false;
bool mIsDefaultMaterial = false;
bool mSpecularAntiAliasing = false;
uint8_t mVariantBits;
// reserve some space to construct the default material instance
mutable FMaterialInstance* mDefaultMaterialInstance = nullptr;
@@ -348,6 +373,13 @@ private:
// current specialization constants for the HwProgram
utils::FixedCapacityVector<backend::Program::SpecializationConstant> mSpecializationConstants;
// Mutable constants defined by this Material.
utils::FixedCapacityVector<MaterialMutableConstant> mMaterialMutableConstants;
// A map from the Constant name to the mMutableMaterialConstant index
std::unordered_map<std::string_view, uint32_t> mMutableSpecializationConstantsNameToIndex;
// Default values for the mutable constants.
backend::Program::MutableSpecConstantsInfo mDefaultMutableSpecializationConstants;
// current push constants for the HwProgram
std::array<utils::FixedCapacityVector<backend::Program::PushConstant>,
backend::Program::SHADER_TYPE_COUNT>
@@ -63,6 +63,7 @@ FMaterialInstance::FMaterialInstance(FEngine& engine, FMaterial const* material,
const char* name) noexcept
: mMaterial(material),
mDescriptorSet("MaterialInstance", material->getDescriptorSetLayout()),
mConstants(material->getDefaultMutableConstants()),
mCulling(CullingMode::BACK),
mShadowCulling(CullingMode::BACK),
mDepthFunc(RasterState::DepthFunc::LE),
@@ -128,6 +129,7 @@ FMaterialInstance::FMaterialInstance(FEngine& engine,
mTextureParameters(other->mTextureParameters),
mDescriptorSet(other->mDescriptorSet.duplicate(
"MaterialInstance", mMaterial->getDescriptorSetLayout())),
mConstants(other->mConstants),
mPolygonOffset(other->mPolygonOffset),
mStencilState(other->mStencilState),
mMaskThreshold(other->mMaskThreshold),
+13 -1
View File
@@ -31,6 +31,7 @@
#include <backend/DriverEnums.h>
#include <backend/Handle.h>
#include <backend/Program.h>
#include <utils/BitmaskEnum.h>
#include <utils/bitset.h>
@@ -66,7 +67,7 @@ public:
void terminate(FEngine& engine);
void commitStreamUniformAssociations(FEngine::DriverApi& driver);
void commit(FEngine& engine) const;
void commit(FEngine::DriverApi& driver) const;
@@ -242,6 +243,10 @@ public:
using MaterialInstance::setParameter;
const backend::Program::MutableSpecConstantsInfo getMutableSpecConstants() const noexcept {
return mConstants;
}
private:
friend class FMaterial;
friend class MaterialInstance;
@@ -264,6 +269,12 @@ private:
template<typename T>
T getParameterImpl(std::string_view name) const;
template<typename T>
void setConstantImpl(std::string_view name, T const& value);
template<typename T>
T getConstantImpl(std::string_view name) const;
// keep these grouped, they're accessed together in the render-loop
FMaterial const* mMaterial = nullptr;
@@ -277,6 +288,7 @@ private:
mutable DescriptorSet mDescriptorSet;
UniformBuffer mUniforms;
bool mHasStreamUniformAssociations = false;
backend::Program::MutableSpecConstantsInfo mConstants;
backend::PolygonOffset mPolygonOffset{};
backend::StencilState mStencilState{};
+1 -1
View File
@@ -442,7 +442,7 @@ void FTexture::setImage(FEngine& engine, size_t const level,
// TODO: PBD should have a p.depth (# layers to skip)
FILAMENT_CHECK_PRECONDITION(
bpp * (pbdWidth - p.left) * (pbdHeight - p.top) * (pbdDepth - 0) >=
bpp * width * height * depth)
bpp * (width - xoffset) * (height - yoffset) * (depth - zoffset))
<< "buffer overflow: (size=" << size_t(p.size) << ", stride=" << size_t(p.stride)
<< ", left=" << unsigned(p.left) << ", top=" << unsigned(p.top)
<< ") smaller than specified region "
+2 -1
View File
@@ -162,7 +162,8 @@ VertexBuffer* VertexBuffer::Builder::build(Engine& engine) {
FILAMENT_CHECK_PRECONDITION(mImpl->mBufferCount <= DOCUMENTED_MAX_VERTEX_BUFFER_COUNT)
<< "bufferCount cannot be more than " << DOCUMENTED_MAX_VERTEX_BUFFER_COUNT;
} else if (mImpl->mBufferCount > DOCUMENTED_MAX_VERTEX_BUFFER_COUNT) {
LOG(WARNING) << "bufferCount cannot be more than " << DOCUMENTED_MAX_VERTEX_BUFFER_COUNT;
utils::slog.w << "bufferCount cannot be more than " << DOCUMENTED_MAX_VERTEX_BUFFER_COUNT
<< utils::io::endl;
}
// Next we check if any unused buffer slots have been allocated. This helps prevent errors
+1 -3
View File
@@ -66,7 +66,6 @@ DescriptorSet& DescriptorSet::operator=(DescriptorSet&& rhs) noexcept {
mDirty = rhs.mDirty;
mValid = rhs.mValid;
mSetAfterCommitWarning = rhs.mSetAfterCommitWarning;
mSetUndefinedParameterWarning = rhs.mSetUndefinedParameterWarning;
}
return *this;
}
@@ -107,12 +106,11 @@ void DescriptorSet::commitSlow(DescriptorSetLayout const& layout,
});
auto const unsetValidDescriptors = layout.getValidDescriptors() & ~mValid;
if (UTILS_VERY_UNLIKELY(!unsetValidDescriptors.empty() && !mSetUndefinedParameterWarning)) {
if (UTILS_VERY_UNLIKELY(!unsetValidDescriptors.empty())) {
unsetValidDescriptors.forEachSetBit([&](auto i) {
LOG(WARNING) << (layout.isSampler(i) ? "Sampler" : "Buffer") << " descriptor " << i
<< " of " << mName.c_str() << " is not set. Please report this issue.";
});
mSetUndefinedParameterWarning = true;
}
}
-1
View File
@@ -110,7 +110,6 @@ private:
mutable utils::bitset64 mValid; // 8
backend::DescriptorSetHandle mDescriptorSetHandle; // 4
mutable bool mSetAfterCommitWarning = false; // 1
mutable bool mSetUndefinedParameterWarning = false; // 1
utils::StaticString mName; // 16
};
+1 -2
View File
@@ -8,8 +8,7 @@ material {
{
type : sampler2d,
name : depth,
precision: high,
unfilterable: true
precision: high
},
{
type : float,
+1 -2
View File
@@ -4,8 +4,7 @@ material {
{
type : sampler2d,
name : depth,
precision: medium,
unfilterable: true
precision: medium
},
{
type : float4,
@@ -9,8 +9,7 @@ material {
{
type : sampler2d,
name : depth,
precision: high,
unfilterable: true
precision: high
},
{
type : float4,
@@ -78,3 +77,4 @@ fragment {
postProcess.color = color;
}
}
+1 -2
View File
@@ -9,8 +9,7 @@ material {
{
type : sampler2d,
name : depth,
precision: medium,
unfilterable: true
precision: medium
},
{
type : float2,
+1 -2
View File
@@ -9,8 +9,7 @@ material {
{
type : sampler2d,
name : depth,
precision: medium,
unfilterable: true
precision: medium
},
{
type : float2,
+1 -2
View File
@@ -5,8 +5,7 @@ material {
type : sampler2d,
name : depth,
precision: high,
multisample : true,
unfilterable: true
multisample : true
}
],
outputs : [
+1 -2
View File
@@ -4,8 +4,7 @@ material {
{
type : sampler2d,
name : depth,
precision: high,
unfilterable: true
precision: high
},
{
type : mat4,
@@ -4,8 +4,7 @@ material {
{
type : sampler2d,
name : depth,
precision: high,
unfilterable: true
precision: high
},
{
type : mat4,
+1 -2
View File
@@ -4,8 +4,7 @@ material {
{
type : sampler2d,
name : depth,
precision: high,
unfilterable: true
precision: high
}
],
variables : [
+1 -2
View File
@@ -4,8 +4,7 @@ material {
{
type : sampler2d,
name : depth,
precision: high,
unfilterable: true
precision: high
},
{
type : mat4,
@@ -4,8 +4,7 @@ material {
{
type : sampler2d,
name : depth,
precision: high,
unfilterable: true
precision: high
},
{
type : mat4,
+2 -2
View File
@@ -1,12 +1,12 @@
Pod::Spec.new do |spec|
spec.name = "Filament"
spec.version = "1.62.0"
spec.version = "1.61.1"
spec.license = { :type => "Apache 2.0", :file => "LICENSE" }
spec.homepage = "https://google.github.io/filament"
spec.authors = "Google LLC."
spec.summary = "Filament is a real-time physically based rendering engine for Android, iOS, Windows, Linux, macOS, and WASM/WebGL."
spec.platform = :ios, "11.0"
spec.source = { :http => "https://github.com/google/filament/releases/download/v1.62.0/filament-v1.62.0-ios.tgz" }
spec.source = { :http => "https://github.com/google/filament/releases/download/v1.61.1/filament-v1.61.1-ios.tgz" }
# Fix linking error with Xcode 12; we do not yet support the simulator on Apple silicon.
spec.pod_target_xcconfig = {
@@ -54,6 +54,7 @@ enum UTILS_PUBLIC ChunkType : uint64_t {
MaterialDescriptorSetLayoutInfo = charTo64bitNum("MAT_DSLI"),
MaterialProperties = charTo64bitNum("MAT_PROP"),
MaterialConstants = charTo64bitNum("MAT_CONS"),
MaterialMutableConstants = charTo64bitNum("MAT_MCNS"),
MaterialPushConstants = charTo64bitNum("MAT_PCON"),
MaterialName = charTo64bitNum("MAT_NAME"),
@@ -28,7 +28,7 @@
namespace filament {
// update this when a new version of filament wouldn't work with older materials
static constexpr size_t MATERIAL_VERSION = 62;
static constexpr size_t MATERIAL_VERSION = 60;
/**
* Supported shading models
@@ -30,7 +30,19 @@ struct MaterialConstant {
ConstantType type;
MaterialConstant() = default;
MaterialConstant(const char* name, ConstantType type) : name(name), type(type) {}
MaterialConstant(utils::CString name, ConstantType type) : name(std::move(name)), type(type) {}
};
struct MaterialMutableConstant {
utils::CString name;
// The Filament engine stores mutable spec constants as a bitfield, meaning that must override
// all of the default values built into in the shader souces. Therefore, unlike the immutable
// constants above, we have to store the default values in the material metadata.
bool defaultValue;
MaterialMutableConstant() = default;
MaterialMutableConstant(utils::CString name, bool defaultValue)
: name(std::move(name)), defaultValue(defaultValue) {}
};
}
@@ -61,7 +61,6 @@ public:
Type type; // type of this sampler
Format format; // format of this sampler
Precision precision; // precision of this sampler
bool unfilterable; // whether the sampling should be unfiltered.
bool multisample; // multisample capable
ShaderStageFlags stages; // stages the sampler can be accessed from
};
@@ -84,7 +83,6 @@ public:
Type type; // type of this sampler
Format format; // format of this sampler
Precision precision; // precision of this sampler
bool unfilterable = false; // whether the sampling should be unfiltered.
bool multisample = false; // multisample capable
ShaderStageFlags stages =
ShaderStageFlags::ALL_SHADER_STAGE_FLAGS; // shader stages using this sampler
@@ -97,8 +95,7 @@ public:
// Add a sampler
Builder& add(std::string_view samplerName, Binding binding, Type type, Format format,
Precision precision = Precision::MEDIUM, bool unfilterable = false,
bool multisample = false,
Precision precision = Precision::MEDIUM, bool multisample = false,
ShaderStageFlags stages = ShaderStageFlags::ALL_SHADER_STAGE_FLAGS) noexcept;
// Add multiple samplers
@@ -50,12 +50,12 @@ SamplerInterfaceBlock::Builder::stageFlags(backend::ShaderStageFlags stageFlags)
}
SamplerInterfaceBlock::Builder& SamplerInterfaceBlock::Builder::add(std::string_view samplerName,
Binding binding, Type type, Format format, Precision precision, bool unfilterable,
bool multisample, ShaderStageFlags stages) noexcept {
Binding binding, Type type, Format format, Precision precision, bool multisample,
ShaderStageFlags stages) noexcept {
mEntries.push_back({
{ samplerName.data(), samplerName.size() }, // name
{}, // uniform name
binding, type, format, precision, unfilterable, multisample, stages });
{ samplerName.data(), samplerName.size() }, // name
{ }, // uniform name
binding, type, format, precision, multisample, stages });
return *this;
}
@@ -65,9 +65,8 @@ SamplerInterfaceBlock SamplerInterfaceBlock::Builder::build() {
SamplerInterfaceBlock::Builder& SamplerInterfaceBlock::Builder::add(
std::initializer_list<ListEntry> list) noexcept {
for (auto& e: list) {
add(e.name, e.binding, e.type, e.format, e.precision, e.unfilterable, e.multisample,
e.stages);
for (auto& e : list) {
add(e.name, e.binding, e.type, e.format, e.precision, e.multisample, e.stages);
}
return *this;
}
+11 -33
View File
@@ -27,7 +27,6 @@
#include <backend/DriverEnums.h>
#include <backend/TargetBufferInfo.h>
#include <mutex>
#include <utils/BitmaskEnum.h>
#include <utils/bitset.h>
#include <utils/compiler.h>
@@ -308,14 +307,15 @@ public:
MaterialBuilder& parameter(const char* name, size_t size, UniformType type,
ParameterPrecision precision = ParameterPrecision::DEFAULT);
//! Add a constant parameter to this material.
//! Add a (mutable) constant parameter to this material.
template<typename T>
using is_supported_constant_parameter_t = typename std::enable_if<
std::is_same<int32_t, T>::value ||
std::is_same<float, T>::value ||
std::is_same<bool, T>::value>::type;
template<typename T, typename = is_supported_constant_parameter_t<T>>
MaterialBuilder& constant(const char *name, ConstantType type, T defaultValue = 0);
MaterialBuilder& constant(const char* name, ConstantType type, bool isMutable,
T defaultValue = 0);
/**
* Add a sampler parameter to this material.
@@ -324,8 +324,8 @@ public:
*/
MaterialBuilder& parameter(const char* name, SamplerType samplerType,
SamplerFormat format = SamplerFormat::FLOAT,
ParameterPrecision precision = ParameterPrecision::DEFAULT, bool unfilterable = false,
bool multisample = false, const char* transformName = "",
ParameterPrecision precision = ParameterPrecision::DEFAULT, bool multisample = false,
const char* transformName = "",
ShaderStageFlags stages = ShaderStageFlags::ALL_SHADER_STAGE_FLAGS);
MaterialBuilder& buffer(filament::BufferInterfaceBlock bib);
@@ -659,39 +659,17 @@ public:
// Sampler
Parameter(const char* paramName, SamplerType t, SamplerFormat f, ParameterPrecision p,
bool unfilterable, bool ms, const char* tn, ShaderStageFlags s)
: name(paramName),
size(1),
precision(p),
samplerType(t),
format(f),
unfilterable(unfilterable),
multisample(ms),
transformName(tn),
stages(s),
parameterType(SAMPLER) {}
bool ms, const char* tn, ShaderStageFlags s)
: name(paramName), size(1), precision(p), samplerType(t), format(f),
parameterType(SAMPLER), multisample(ms), transformName(tn), stages(s) { }
// Uniform
Parameter(const char* paramName, UniformType t, size_t typeSize, ParameterPrecision p)
: name(paramName),
size(typeSize),
uniformType(t),
precision(p),
format{ 0 },
unfilterable(false),
multisample(false),
parameterType(UNIFORM) {}
: name(paramName), size(typeSize), uniformType(t), precision(p), parameterType(UNIFORM) { }
// Subpass
Parameter(const char* paramName, SubpassType t, SamplerFormat f, ParameterPrecision p)
: name(paramName),
size(1),
precision(p),
subpassType(t),
format(f),
unfilterable(false),
multisample(false),
parameterType(SUBPASS) {}
: name(paramName), size(1), precision(p), subpassType(t), format(f), parameterType(SUBPASS) { }
utils::CString name;
size_t size;
@@ -700,7 +678,6 @@ public:
SamplerType samplerType;
SubpassType subpassType;
SamplerFormat format;
bool unfilterable;
bool multisample;
utils::CString transformName;
ShaderStageFlags stages;
@@ -895,6 +872,7 @@ private:
PropertyList mProperties;
ParameterList mParameters;
ConstantList mConstants;
ConstantList mMutableConstants;
PushConstantList mPushConstants;
SubpassList mSubpasses;
VariableList mVariables;
+39 -24
View File
@@ -300,29 +300,36 @@ MaterialBuilder& MaterialBuilder::parameter(const char* name, UniformType const
return parameter(name, 1, type, precision);
}
MaterialBuilder& MaterialBuilder::parameter(const char* name, SamplerType samplerType,
SamplerFormat format, ParameterPrecision precision, bool unfilterable, bool multisample,
SamplerFormat format, ParameterPrecision precision, bool multisample,
const char* transformName, ShaderStageFlags stages) {
FILAMENT_CHECK_PRECONDITION(
!multisample || (format != SamplerFormat::SHADOW &&
(samplerType == SamplerType::SAMPLER_2D ||
samplerType == SamplerType::SAMPLER_2D_ARRAY)))
FILAMENT_CHECK_PRECONDITION(!multisample ||
(format != SamplerFormat::SHADOW &&
(samplerType == SamplerType::SAMPLER_2D ||
samplerType == SamplerType::SAMPLER_2D_ARRAY)))
<< "multisample samplers only possible with SAMPLER_2D or SAMPLER_2D_ARRAY,"
" as long as type is not SHADOW";
FILAMENT_CHECK_POSTCONDITION(mParameterCount < MAX_PARAMETERS_COUNT) << "Too many parameters";
mParameters[mParameterCount++] = { name, samplerType, format, precision, unfilterable,
multisample, transformName, stages };
mParameters[mParameterCount++] = { name, samplerType, format, precision, multisample, transformName, stages };
return *this;
}
template<typename T, typename>
MaterialBuilder& MaterialBuilder::constant(const char* name, ConstantType const type, T defaultValue) {
auto result = std::find_if(mConstants.begin(), mConstants.end(), [name](const Constant& c) {
return c.name == CString(name);
});
FILAMENT_CHECK_POSTCONDITION(result == mConstants.end())
MaterialBuilder& MaterialBuilder::constant(const char* name, ConstantType const type,
bool isMutable, T defaultValue) {
ConstantList& constants = isMutable ? mMutableConstants : mConstants;
const ConstantList& otherConstants = isMutable ? mConstants : mMutableConstants;
auto result = std::find_if(constants.begin(), constants.end(),
[name](const Constant& c) { return c.name == CString(name); });
FILAMENT_CHECK_POSTCONDITION(result == constants.end())
<< "There is already a constant parameter present with the name " << name << ".";
auto otherResult = std::find_if(otherConstants.begin(), otherConstants.end(),
[name](const Constant& c) { return c.name == CString(name); });
FILAMENT_CHECK_POSTCONDITION(otherResult == otherConstants.end())
<< "There is already a constant parameter present with the name " << name << ".";
Constant constant {
.name = CString(name),
.type = type,
@@ -354,15 +361,20 @@ MaterialBuilder& MaterialBuilder::constant(const char* name, ConstantType const
assert_invariant(false);
}
mConstants.push_back(constant);
if (isMutable) {
FILAMENT_CHECK_POSTCONDITION(type == ConstantType::BOOL)
<< "Mutable spec constants must be bool type.";
}
constants.push_back(constant);
return *this;
}
template MaterialBuilder& MaterialBuilder::constant<int32_t>(
const char* name, ConstantType type, int32_t defaultValue);
const char* name, ConstantType type, bool isMutable, int32_t defaultValue);
template MaterialBuilder& MaterialBuilder::constant<float>(
const char* name, ConstantType type, float defaultValue);
const char* name, ConstantType type, bool isMutable, float defaultValue);
template MaterialBuilder& MaterialBuilder::constant<bool>(
const char* name, ConstantType type, bool defaultValue);
const char* name, ConstantType type, bool isMutable, bool defaultValue);
MaterialBuilder& MaterialBuilder::buffer(BufferInterfaceBlock bib) {
FILAMENT_CHECK_POSTCONDITION(mBuffers.size() < MAX_BUFFERS_COUNT) << "Too many buffers";
@@ -639,12 +651,10 @@ void MaterialBuilder::prepareToBuild(MaterialInfo& info) noexcept {
assert_invariant(!param.isSubpass());
if (param.isSampler()) {
sbb.add({ param.name.data(), param.name.size() }, binding, param.samplerType,
param.format, param.precision, param.unfilterable, param.multisample,
param.stages);
param.format, param.precision, param.multisample, param.stages);
if (!param.transformName.empty()) {
ibb.add({ { { param.transformName.data(), param.transformName.size() },
uint8_t(binding), 0, UniformType::MAT3, Precision::DEFAULT,
FeatureLevel::FEATURE_LEVEL_0 } });
ibb.add({{{ param.transformName.data(), param.transformName.size() }, uint8_t(binding),
0, UniformType::MAT3, Precision::DEFAULT, FeatureLevel::FEATURE_LEVEL_0 }});
}
binding++;
} else if (param.isUniform()) {
@@ -909,7 +919,7 @@ bool MaterialBuilder::generateShaders(JobSystem& jobSystem, const std::vector<Va
BlobDictionary spirvDictionary;
// End: must be protected by lock
ShaderGenerator sg(mProperties, mVariables, mOutputs, mDefines, mConstants, mPushConstants,
ShaderGenerator sg(mProperties, mVariables, mOutputs, mDefines, mConstants, mMutableConstants, mPushConstants,
mMaterialFragmentCode.getResolved(), mMaterialFragmentCode.getLineOffset(),
mMaterialVertexCode.getResolved(), mMaterialVertexCode.getLineOffset(),
mMaterialDomain);
@@ -1494,7 +1504,7 @@ bool MaterialBuilder::needsStandardDepthProgram() const noexcept {
std::string MaterialBuilder::peek(backend::ShaderStage const stage,
const CodeGenParams& params, const PropertyList& properties) noexcept {
ShaderGenerator const sg(properties, mVariables, mOutputs, mDefines, mConstants, mPushConstants,
ShaderGenerator const sg(properties, mVariables, mOutputs, mDefines, mConstants, mMutableConstants, mPushConstants,
mMaterialFragmentCode.getResolved(), mMaterialFragmentCode.getLineOffset(),
mMaterialVertexCode.getResolved(), mMaterialVertexCode.getLineOffset(),
mMaterialDomain);
@@ -1620,9 +1630,14 @@ void MaterialBuilder::writeCommonChunks(ChunkContainer& container, MaterialInfo&
// User constant parameters
FixedCapacityVector<MaterialConstant> constantsEntry(mConstants.size());
std::transform(mConstants.begin(), mConstants.end(), constantsEntry.begin(),
[](Constant const& c) { return MaterialConstant(c.name.c_str(), c.type); });
[](Constant const& c) { return MaterialConstant(c.name, c.type); });
container.push<MaterialConstantParametersChunk>(std::move(constantsEntry));
FixedCapacityVector<MaterialMutableConstant> mutableConstantsEntry(mMutableConstants.size());
std::transform(mMutableConstants.begin(), mMutableConstants.end(), mutableConstantsEntry.begin(),
[](Constant const& c) { return MaterialMutableConstant(c.name, c.defaultValue.b); });
container.push<MaterialMutableConstantParametersChunk>(std::move(mutableConstantsEntry));
FixedCapacityVector<MaterialPushConstant> pushConstantsEntry(mPushConstants.size());
std::transform(mPushConstants.begin(), mPushConstants.end(), pushConstantsEntry.begin(),
[](PushConstant const& c) {
@@ -77,7 +77,6 @@ void MaterialSamplerInterfaceBlockChunk::flatten(Flattener& f) {
f.writeUint8(static_cast<uint8_t>(sInfo.type));
f.writeUint8(static_cast<uint8_t>(sInfo.format));
f.writeUint8(static_cast<uint8_t>(sInfo.precision));
f.writeBool(sInfo.unfilterable);
f.writeBool(sInfo.multisample);
}
}
@@ -118,6 +117,20 @@ void MaterialConstantParametersChunk::flatten(Flattener& f) {
// ------------------------------------------------------------------------------------------------
MaterialMutableConstantParametersChunk::MaterialMutableConstantParametersChunk(
FixedCapacityVector<MaterialMutableConstant> constants)
: Chunk(MaterialMutableConstants), mConstants(std::move(constants)) {}
void MaterialMutableConstantParametersChunk::flatten(Flattener& f) {
f.writeUint64(mConstants.size());
for (const auto& constant : mConstants) {
f.writeString(constant.name.c_str());
f.writeBool(constant.defaultValue);
}
}
// ------------------------------------------------------------------------------------------------
MaterialPushConstantParametersChunk::MaterialPushConstantParametersChunk(
CString const& structVarName, FixedCapacityVector<MaterialPushConstant> constants)
: Chunk(MaterialPushConstants),
@@ -231,11 +244,7 @@ void MaterialDescriptorSetLayoutChunk::flatten(Flattener& f) {
f.writeUint8(uint8_t(descriptor_sets::getDescriptorType(entry.type, entry.format)));
f.writeUint8(uint8_t(entry.stages));
f.writeUint8(entry.binding);
if (entry.unfilterable) {
f.writeUint8(uint8_t(DescriptorFlags::UNFILTERABLE));
} else {
f.writeUint8(uint8_t(DescriptorFlags::NONE));
}
f.writeUint8(uint8_t(DescriptorFlags::NONE));
f.writeUint16(0);
}
}
@@ -18,6 +18,7 @@
#define TNT_FILAMAT_MAT_INTEFFACE_BLOCK_CHUNK_H
#include "Chunk.h"
#include "private/filament/ConstantInfo.h"
#include <backend/Program.h>
@@ -91,6 +92,20 @@ private:
// ------------------------------------------------------------------------------------------------
class MaterialMutableConstantParametersChunk final : public Chunk {
public:
explicit MaterialMutableConstantParametersChunk(
FixedCapacityVector<filament::MaterialMutableConstant> constants);
~MaterialMutableConstantParametersChunk() override = default;
private:
void flatten(Flattener&) override;
FixedCapacityVector<filament::MaterialMutableConstant> mConstants;
};
// ------------------------------------------------------------------------------------------------
class MaterialPushConstantParametersChunk final : public Chunk {
public:
explicit MaterialPushConstantParametersChunk(CString const& structVarName,
+17 -8
View File
@@ -283,10 +283,17 @@ void ShaderGenerator::appendShader(io::sstream& ss,
}
}
void ShaderGenerator::generateUserSpecConstants(
const CodeGenerator& cg, io::sstream& fs, MaterialBuilder::ConstantList const& constants) {
void ShaderGenerator::generateAllUserSpecConstants(const CodeGenerator& cg, io::sstream& fs,
MaterialBuilder::ConstantList const& constants,
MaterialBuilder::ConstantList const& mutableConstants) {
generateUserSpecConstants(cg, fs, constants, 0);
generateUserSpecConstants(cg, fs, mutableConstants, constants.size());
}
void ShaderGenerator::generateUserSpecConstants(const CodeGenerator& cg, io::sstream& fs,
MaterialBuilder::ConstantList const& constants, size_t offset) {
// Constants 0 to CONFIG_MAX_RESERVED_SPEC_CONSTANTS - 1 are reserved by Filament.
size_t index = CONFIG_MAX_RESERVED_SPEC_CONSTANTS;
size_t index = CONFIG_MAX_RESERVED_SPEC_CONSTANTS + offset;
for (const auto& constant : constants) {
std::string const fullName = std::string("materialConstants_") + constant.name.c_str();
switch (constant.type) {
@@ -314,6 +321,7 @@ ShaderGenerator::ShaderGenerator(
MaterialBuilder::OutputList const& outputs,
MaterialBuilder::PreprocessorDefineList const& defines,
MaterialBuilder::ConstantList const& constants,
MaterialBuilder::ConstantList const& mutableConstants,
MaterialBuilder::PushConstantList const& pushConstants,
CString const& materialCode, size_t const lineOffset,
CString const& materialVertexCode, size_t const vertexLineOffset,
@@ -336,6 +344,7 @@ ShaderGenerator::ShaderGenerator(
mMaterialDomain = materialDomain;
mDefines = defines;
mConstants = constants;
mMutableConstants = mutableConstants;
mPushConstants = pushConstants;
if (mMaterialFragmentCode.empty()) {
@@ -392,7 +401,7 @@ std::string ShaderGenerator::createSurfaceVertexProgram(ShaderModel const shader
cg.generateCommonProlog(vs, ShaderStage::VERTEX, material, variant);
generateUserSpecConstants(cg, vs, mConstants);
generateAllUserSpecConstants(cg, vs, mConstants, mMutableConstants);
// note: even if the user vertex shader is empty, we can't use the "optimized" version if
// we're in masked mode because fragment shader needs the color varyings
@@ -525,7 +534,7 @@ std::string ShaderGenerator::createSurfaceFragmentProgram(ShaderModel const shad
io::sstream fs;
cg.generateCommonProlog(fs, ShaderStage::FRAGMENT, material, variant);
generateUserSpecConstants(cg, fs, mConstants);
generateAllUserSpecConstants(cg, fs, mConstants, mMutableConstants);
generateSurfaceMaterialVariantDefines(
fs, ShaderStage::FRAGMENT, featureLevel, material, variant);
@@ -681,7 +690,7 @@ std::string ShaderGenerator::createSurfaceComputeProgram(ShaderModel const shade
cg.generateCommonProlog(s, ShaderStage::COMPUTE, material, {});
generateUserSpecConstants(cg, s, mConstants);
generateAllUserSpecConstants(cg, s, mConstants, mMutableConstants);
CodeGenerator::generateSurfaceTypes(s, ShaderStage::COMPUTE);
@@ -721,7 +730,7 @@ std::string ShaderGenerator::createPostProcessVertexProgram(ShaderModel const sm
io::sstream vs;
cg.generateCommonProlog(vs, ShaderStage::VERTEX, material, {});
generateUserSpecConstants(cg, vs, mConstants);
generateAllUserSpecConstants(cg, vs, mConstants, mMutableConstants);
CodeGenerator::generateDefine(vs, "LOCATION_POSITION", uint32_t(POSITION));
@@ -765,7 +774,7 @@ std::string ShaderGenerator::createPostProcessFragmentProgram(ShaderModel const
io::sstream fs;
cg.generateCommonProlog(fs, ShaderStage::FRAGMENT, material, {});
generateUserSpecConstants(cg, fs, mConstants);
generateAllUserSpecConstants(cg, fs, mConstants, mMutableConstants);
generatePostProcessMaterialVariantDefines(fs, PostProcessVariant(variant));
+8 -1
View File
@@ -51,6 +51,7 @@ public:
MaterialBuilder::OutputList const& outputs,
MaterialBuilder::PreprocessorDefineList const& defines,
MaterialBuilder::ConstantList const& constants,
MaterialBuilder::ConstantList const& mutableConstants,
MaterialBuilder::PushConstantList const& pushConstants,
utils::CString const& materialCode,
size_t lineOffset,
@@ -104,9 +105,14 @@ private:
static void generatePostProcessMaterialVariantDefines(utils::io::sstream& out,
filament::PostProcessVariant variant) noexcept;
static void generateAllUserSpecConstants(const CodeGenerator& cg, utils::io::sstream& fs,
MaterialBuilder::ConstantList const& constants,
MaterialBuilder::ConstantList const& mutableConstants);
static void generateUserSpecConstants(
const CodeGenerator& cg, utils::io::sstream& fs,
MaterialBuilder::ConstantList const& constants);
MaterialBuilder::ConstantList const& constants,
size_t offset);
std::string createPostProcessVertexProgram(filament::backend::ShaderModel sm,
MaterialBuilder::TargetApi targetApi, MaterialBuilder::TargetLanguage targetLanguage,
@@ -135,6 +141,7 @@ private:
MaterialBuilder::MaterialDomain mMaterialDomain;
MaterialBuilder::PreprocessorDefineList mDefines;
MaterialBuilder::ConstantList mConstants;
MaterialBuilder::ConstantList mMutableConstants;
MaterialBuilder::PushConstantList mPushConstants;
utils::CString mMaterialFragmentCode; // fragment or compute code
utils::CString mMaterialVertexCode;
+7 -7
View File
@@ -888,10 +888,10 @@ TEST_F(MaterialCompiler, ConstantParameter) {
}
)");
filamat::MaterialBuilder builder;
builder.constant("myFloatConstant", ConstantType::FLOAT, 1.0f);
builder.constant("myIntConstant", ConstantType::INT, 123);
builder.constant("myBoolConstant", ConstantType::BOOL, true);
builder.constant<bool>("myOtherBoolConstant", ConstantType::BOOL);
builder.constant("myFloatConstant", ConstantType::FLOAT, /*isMutable=*/false, 1.0f);
builder.constant("myIntConstant", ConstantType::INT, /*isMutable=*/false, 123);
builder.constant("myBoolConstant", ConstantType::BOOL, /*isMutable=*/false, true);
builder.constant<bool>("myOtherBoolConstant", ConstantType::BOOL, /*isMutable=*/false);
builder.shading(filament::Shading::LIT);
builder.material(shaderCode.c_str());
@@ -904,8 +904,8 @@ TEST_F(MaterialCompiler, ConstantParameterSameName) {
#ifdef __EXCEPTIONS
EXPECT_THROW({
filamat::MaterialBuilder builder;
builder.constant("myFloatConstant", ConstantType::FLOAT, 1.0f);
builder.constant("myFloatConstant", ConstantType::FLOAT, 1.0f);
builder.constant("myFloatConstant", ConstantType::FLOAT, /*isMutable=*/false, 1.0f);
builder.constant("myFloatConstant", ConstantType::FLOAT, /*isMutable=*/false, 1.0f);
}, utils::PostconditionPanic);
#endif
}
@@ -914,7 +914,7 @@ TEST_F(MaterialCompiler, ConstantParameterWrongType) {
#ifdef __EXCEPTIONS
EXPECT_THROW({
filamat::MaterialBuilder builder;
builder.constant("myFloatConstant", ConstantType::FLOAT, 10);
builder.constant("myFloatConstant", ConstantType::FLOAT, /*isMutable=*/false, 10);
}, utils::PostconditionPanic);
#endif
}
@@ -21,7 +21,6 @@
#include <filament/Material.h>
#include <filament/MaterialInstance.h>
#include <utils/FixedCapacityVector.h>
#include <utils/compiler.h>
#include <array>
@@ -197,7 +196,6 @@ void processShaderString(std::string* shader, const UvMap& uvmap,
* Creates a material provider that builds materials on the fly, composing GLSL at run time.
*
* @param optimizeShaders Optimizes shaders, but at significant cost to construction time.
* @param variantFilters Filter out variants that are not required.
* @return New material provider that can build materials at run time.
*
* Requires \c libfilamat to be linked in. Not available in \c libgltfio_core.
@@ -205,8 +203,7 @@ void processShaderString(std::string* shader, const UvMap& uvmap,
* @see createUbershaderProvider
*/
UTILS_PUBLIC
MaterialProvider* createJitShaderProvider(Engine* engine, bool optimizeShaders = false,
utils::FixedCapacityVector<char const*> const& variantFilters = {});
MaterialProvider* createJitShaderProvider(Engine* engine, bool optimizeShaders = false);
/**
* Creates a material provider that loads a small set of pre-built materials.

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