Initial commit

.gitignore

@@ -0,0 +1,11 @@
+/keys.store
+/build
+debug
+__debug_bin
+R.java
+*.env
+.DS_Store
+*.jar
+*.aar
+*.[xX]cframework
+*.[fF]ramework

.vscode/launch.json

@@ -0,0 +1,33 @@
+{
+    // Use IntelliSense to learn about possible attributes.
+    // Hover to view descriptions of existing attributes.
+    // For more information, visit: https://go.microsoft.com/fwlink/?linkid=830387
+    "version": "0.2.0",
+    "configurations": [
+        {
+            "name": "In integrated terminal. Breaks bluetooth",
+            "type": "go",
+            "request": "launch",
+            "cwd": "${workspaceFolder}",
+            "program": "./desktop"
+        },
+        {
+            "name": "In separate Mac OS Terminal with command: dlv dap --listen :6868",
+            "type": "go",
+            "request": "launch",
+            "cwd": "${workspaceFolder}",
+            "envFile": "${workspaceFolder}/.env",
+            "program": "${workspaceFolder}/desktop",
+            "port": 6868
+        },
+        {
+            "name": "Remote device",
+            "type": "go",
+            "request": "attach",
+            "mode": "remote",
+            "remotePath": "${workspaceFolder}",
+            "port": 4334,
+            "host": "127.0.0.1",
+        }
+    ]
+}

.vscode/settings.json

@@ -0,0 +1,12 @@
+{
+    "java.project.referencedLibraries": {
+
+        "include": [
+            "build/jars/**/*.jar",
+        ]
+    },
+    "java.jdt.ls.vmargs": "-XX:+UseParallelGC -XX:GCTimeRatio=4 -XX:AdaptiveSizePolicyWeight=90 -Dsun.zip.disableMemoryMapping=true -Xmx8G -Xms100m",
+    "java.project.sourcePaths": [
+        "android/java"
+    ]
+}

.vscode/tasks.json

@@ -0,0 +1,21 @@
+{
+    // See https://go.microsoft.com/fwlink/?LinkId=733558
+    // for the documentation about the tasks.json format
+    "version": "2.0.0",
+    "tasks": [
+        {
+            "label": "makeDebug",
+            "type": "shell",
+            "command": "make debug",
+        },
+        {
+            "label": "build",
+            "type": "shell",
+            "group": {
+                "kind": "build",
+                "isDefault": true
+            },
+            "command": "go build -o debug github.com/empathicqubit/giouibind/desktop"
+        }
+    ]
+}

Makefile

@@ -0,0 +1,178 @@
+# FIXME Explain wth we're doing
+MIN_SDK_VERSION ?= 21
+ANDROID_SDK_VERSION ?= 30.0.3
+ANDROID_HOME ?= $(HOME)/Android/Sdk
+ANDROID_NDK_HOME ?= $(ANDROID_HOME)/ndk/24.0.8215888
+MAX_SDK_VERSION ?= 32
+
+GDB_PORT ?= 4334
+
+# Target package name.
+ANDROID_PACKAGE ?= gl.entan.giouibind
+
+# Settings for keystore which is used to sign APK.
+KEYS_DN=DC=gl,CN=entan
+KEYS_PASS=123456
+KEYS_VALIDITY=365
+KEYS_ALGORITHM=RSA
+KEYS_SIZE=2048
+
+# Target build dir. Resulting APK file will be available here.
+BUILD_DIR=build
+
+_BUILD_TOOLS=$(ANDROID_HOME)/build-tools/$(ANDROID_SDK_VERSION)
+_ANDROID_JAR_PATH=$(ANDROID_HOME)/platforms/android-$(MAX_SDK_VERSION)/android.jar
+_SWIFT_SRC=$(shell find ios -iname '*.swift')
+_JAVA_SRC=$(shell find android/java -iname '*.java')
+_GO_SRC=$(shell find . -iname '*.go')
+_JAVA_ROOT_PATH=android/java/$(subst .,/,$(ANDROID_PACKAGE))
+_ADB_PATH=$(ANDROID_HOME)/platform-tools/adb
+_JARS=build/jars/ble/classes.jar build/jars/go-android/classes.jar
+
+
+build: build-ios build-android
+
+install-ios: build/ios-app.ipa
+	ideviceinstaller -i build/ios-app.ipa/giouibind.ipa
+
+ios: build-ios
+build-ios: build/ios-app.ipa
+
+android: build-android
+build-android: build/android-app.apk
+
+run-android: install-android
+	$(_ADB_PATH) shell am start -n $(ANDROID_PACKAGE)/.MainActivity
+	$(_ADB_PATH) shell 'while ! dumpsys window windows | grep -o "$(ANDROID_PACKAGE)" 2>&1 > /dev/null ; do sleep 1 ; done'
+
+debug-android: run-android
+#FIXME Delve
+	$(_ADB_PATH) push $(ANDROID_NDK_HOME)/prebuilt/android-arm64/gdbserver/gdbserver /data/local/tmp
+	$(_ADB_PATH) shell "chmod 777 /data/local/tmp/gdbserver"
+	$(_ADB_PATH) forward tcp:$(GDB_PORT) tcp:$(GDB_PORT)
+	$(_ADB_PATH) shell 'su -c killall gdbserver || exit 0'
+	$(_ADB_PATH) shell 'su -c set enforce 0'
+	$(_ADB_PATH) shell 'su -c /data/local/tmp/gdbserver :$(GDB_PORT) --attach $$(ps -A -o NAME,PID | grep "$(ANDROID_PACKAGE)" | cut -F 2)'
+
+install-android: build-android
+	$(_ADB_PATH) install build/android-app.apk
+
+# Initialize keystore to sign APK.
+keys.store:
+	keytool -genkeypair \
+		-validity $(KEYS_VALIDITY) \
+		-keystore $@ \
+		-keyalg $(KEYS_ALGORITHM) \
+		-keysize $(KEYS_SIZE) \
+		-storepass $(KEYS_PASS) \
+		-keypass $(KEYS_PASS) \
+		-dname $(KEYS_DN) \
+		-deststoretype pkcs12
+
+build/Mobile.xcframework: $(_GO_SRC)
+	CGO_ENABLED=1 GO386=softfloat gomobile bind \
+		-target ios \
+		-o "$@" \
+		github.com/empathicqubit/giouibind/mobile
+
+build/ios-app.ipa: build/ios.xcarchive
+	xcodebuild -allowProvisioningUpdates -exportArchive -archivePath build/ios.xcarchive -exportOptionsPlist ios/export-options.plist -exportPath "$@"
+
+build/ios.xcarchive: $(_SWIFT_SRC) build/Mobile.xcframework
+	xcodebuild -project ios/giouibind/giouibind.xcodeproj -scheme giouibind -sdk iphoneos -configuration AppStoreDistribution archive -archivePath "$@"
+
+build/jars/ble/classes.jar:
+	@mkdir -p build/jars/ble
+
+	curl -L -o "build/ble.aar" https://repo1.maven.org/maven2/no/nordicsemi/android/ble/2.4.0/ble-2.4.0.aar
+	cd build/jars/ble && unzip -o ../../ble.aar
+	touch "$@"
+
+build/jars/go-android/classes.jar: $(_GO_SRC) $(ANDROID_NDK_HOME)
+	@mkdir -p build/jars/go-android
+
+	CGO_ENABLED=1 ANDROID_NDK_HOME=$(ANDROID_NDK_HOME) GO386=softfloat gomobile bind \
+		-target android \
+		-javapkg $(ANDROID_PACKAGE) \
+		-o "build/go-android.aar" \
+		github.com/empathicqubit/giouibind/mobile
+
+	# Unpack resulting AAR library to link it to APK during further stages.
+	@unzip -o -qq "build/go-android.aar" -d build/jars/go-android
+	@ln -sf jars/go-android/jni build/lib
+	touch "$@"
+
+# Collect resources and generate R.java.
+$(_JAVA_ROOT_PATH)/R.java: $(wildcard android/res/*/*.*) android/AndroidManifest.xml $(_ANDROID_JAR_PATH)
+	$(_BUILD_TOOLS)/aapt package \
+		-f \
+		-m \
+		-J android/java \
+		-M android/AndroidManifest.xml \
+		-S android/res \
+		-I $(_ANDROID_JAR_PATH)
+
+# Generate a JAR suitable for code completion (less java.* classes)
+build/jars/android-meta.jar: $(_ANDROID_JAR_PATH)
+	@mkdir -p build/jars
+
+	cp "$(_ANDROID_JAR_PATH)" "$@"
+	zip -d "$@" 'java/*'
+
+build/obj.jar: build/jars/android-meta.jar $(_JARS) $(_JAVA_SRC) $(_JAVA_ROOT_PATH)/R.java
+	@mkdir -p build/obj
+
+	javac \
+		-source 8 \
+		-target 8 \
+		-d build/obj \
+		-classpath $(_ANDROID_JAR_PATH):android/java:$(subst $() $(),:,$(_JARS)) \
+		$(_JAVA_SRC)
+	
+	jar cvf "$@" -C build/obj/ .
+
+# Convert compiled Java code into DEX file (required by Android).
+build/classes.dex: build/d8.jar
+	$(_BUILD_TOOLS)/dx \
+		--dex \
+		--min-sdk-version $(MIN_SDK_VERSION) \
+		--output build/classes.dex \
+		build/d8.jar
+
+build/d8.jar: build/obj.jar
+	$(_BUILD_TOOLS)/d8 \
+		--output build/d8.jar \
+		--classpath $(_ANDROID_JAR_PATH) \
+		$(_JARS) \
+		build/obj.jar
+
+# Package everything into unaligned APK file.
+build/app.apk.unaligned: build/classes.dex
+	$(_BUILD_TOOLS)/aapt package \
+		-f \
+		-m \
+		-F build/app.apk.unaligned \
+		-M android/AndroidManifest.xml \
+		-S android/res \
+		-I $(_ANDROID_JAR_PATH)
+
+	cd build && $(_BUILD_TOOLS)/aapt add \
+		app.apk.unaligned \
+		classes.dex \
+		lib/*/*
+
+# Align unaligned APK file and sign it using keystore.
+build/android-app.apk: keys.store build/app.apk.unaligned
+	$(_BUILD_TOOLS)/zipalign \
+		-f 4 \
+		build/app.apk.unaligned \
+		"$@"
+
+	$(_BUILD_TOOLS)/apksigner sign \
+		--ks keys.store \
+		--ks-pass pass:$(KEYS_PASS) \
+		"$@"
+
+clean:
+	@rm -rf build
+	@rm -rf $(_JAVA_ROOT_PATH)/R.java

README.md

@@ -0,0 +1,49 @@
+# Minimal Gio-Powered Android/iOS/desktop app
+
+This project serves as an example how to build minimal working Gio-powered
+Android/iOS/desktop application **without** Android Studio or Gradle. The example
+was intended to connect to a BLE device with a specific service identifier.
+Please search for FIXME in the project and replace information as appropriate.
+Alternatively you can remove this stuff and replace it with your own.
+
+The interface INativeBridge in native/native.go contains the methods which
+bridge go to either Java on Android, ObjC/Swift on iOS, or more Go classes
+on desktop.
+
+Makefile targets:
+* **build**: Build for both iOS and Android. Only works on Mac OS because of iOS
+* **build-ios**, **build-android**: Build for single platform
+* **install-ios**, **install-android**: Install to a real device. Your signing
+team ID in /ios/export-options.plist must be correct.
+* **run-android**: Tells the application to start on Android
+* **debug-android**: Doesn't work yet.
+
+## Requirements
+
+To be able to build, following dependencies are
+required:
+
+* Android SDK version 32, be sure to set ANDROID\_HOME environment variable
+* Android NDK version 24.0.8215888
+* Android build-tools version 30.0.3
+* Java on your PATH
+* gomobile `go install "golang.org/x/mobile/cmd/gomobile"`
+
+## Description of different project files
+* **build**: All the build stuff goes here
+* **android**: All the Android-specific files here. If you want to include
+images in your application, it's probably better to do that in Go 
+with //go:embed, instead of including images in the res folder. GodObject
+is the main class which implements the bridge with Go. The Gio classes are
+overridden to work around some issues with using Gio with `gomobile bind`
+* **ios**: XCode project which contains iOS specific INativeBridge implementation.
+* **mobile**: The entrypoint for `gomobile bind`
+* **mobile/gio**: A stub entrypoint for Gio. You shouldn't need to change this.
+* **desktop**: Entrypoint for the desktop application. The example only works
+on Mac OS because the cbgo library is made for Mac OS bluetooth only, but you
+can replace it with whatever you want and that should work on MacOS/Linux/Windows.
+* **gio**: Stubs for Gio packages which fail to build on Linux when targeting Android.
+
+## Credits
+
+Thanks to [seletskiy/ebiten-android-minimal](https://github.com/seletskiy/ebiten-android-minimal) and [GioUI](https://gioui.org/)

android/AndroidManifest.xml

@@ -0,0 +1,25 @@
+<?xml version='1.0'?>
+<manifest xmlns:android='http://schemas.android.com/apk/res/android' package='gl.entan.giouibind' android:versionCode='0' android:versionName='0'>
+    <uses-sdk android:minSdkVersion="21"
+        android:targetSdkVersion="29"
+        android:maxSdkVersion="32" />
+
+    <uses-feature android:glEsVersion="0x00030000"/>
+
+    <uses-permission android:name="android.permission.BLUETOOTH" />
+    <uses-permission android:name="android.permission.ACCESS_BACKGROUND_LOCATION" />
+
+    <application android:icon="@drawable/icon" android:label='TestApp'>
+        <activity 
+            android:icon="@drawable/icon"
+            android:name='.MainActivity'
+            android:theme="@android:style/Theme.NoTitleBar.Fullscreen"
+            android:configChanges="orientation|keyboardHidden"
+            android:windowSoftInputMode="adjustResize">
+             <intent-filter>
+                <category android:name='android.intent.category.LAUNCHER'/>
+                <action android:name='android.intent.action.MAIN'/>
+             </intent-filter>
+        </activity>
+    </application>
+</manifest>

android/java/gl/entan/giouibind/GodObject.java

@@ -0,0 +1,108 @@
+package gl.entan.giouibind;
+
+import java.util.ArrayList;
+import java.util.List;
+
+import android.Manifest.permission;
+import android.app.Activity;
+import android.bluetooth.BluetoothAdapter;
+import android.bluetooth.BluetoothDevice;
+import android.bluetooth.BluetoothProfile;
+import android.bluetooth.BluetoothSocket;
+import android.content.Context;
+import android.content.pm.PackageManager;
+import android.util.Log;
+import gl.entan.giouibind.mobile.IGodObject;
+import gl.entan.giouibind.mobile.Mobile;
+
+public class GodObject implements IGodObject {
+    private static GodObject instance = null;
+
+    private BluetoothSocket socket = null;
+    private Context context = null;
+    private List<MyBleManager> managers;
+
+    private GodObject(Context context) {
+        this.context = context;
+        this.managers = new ArrayList<>();
+    }
+
+    public static GodObject getGodObject(Context context) {
+        if(GodObject.instance == null) {
+            GodObject.instance = new GodObject(context);
+        }
+
+        return GodObject.instance;
+    }
+    
+    public void disconnectFromDevice() {
+        if(this.socket != null) {
+            try {
+                this.socket.close();
+            }
+            catch (Exception e) {
+                Log.e("bananas", "Couldn't close socket", e);
+            }
+        }
+    }
+
+    public void connectToDevice() {
+        GodObject self = this;
+
+        this.disconnectFromDevice();
+
+        BluetoothAdapter.getDefaultAdapter().getProfileProxy(this.context, new BluetoothProfile.ServiceListener() {
+            @Override
+            public void onServiceDisconnected(int arg0) {
+            }
+
+            @Override
+            public void onServiceConnected(int profile, BluetoothProfile proxy) {
+                Log.i("bananas", "Called service connected");
+
+                for (BluetoothDevice device : proxy.getConnectedDevices()) {
+                    MyBleManager manager = new MyBleManager(context);
+
+                    manager
+                        .connect(device)
+                        .retry(20)
+                        .timeout(5000)
+                        .useAutoConnect(true)
+                        .enqueue();
+
+                    self.managers.add(manager);
+                }
+
+                BluetoothAdapter.getDefaultAdapter().closeProfileProxy(profile, proxy);
+            }
+        }, BluetoothProfile.HEADSET);
+    }
+
+    @Override
+    public boolean enableBluetooth() {
+        Context context = this.context;
+
+        if(context.checkSelfPermission(permission.BLUETOOTH) == PackageManager.PERMISSION_DENIED) {
+            ((Activity)context).requestPermissions(new String[] { permission.BLUETOOTH }, 0);
+            return false;
+        }
+
+        if(context.checkSelfPermission(permission.ACCESS_BACKGROUND_LOCATION) == PackageManager.PERMISSION_DENIED) {
+            ((Activity)context).requestPermissions(new String[] { permission.ACCESS_BACKGROUND_LOCATION }, 0);
+            return false;
+        }
+
+        return true;
+    }
+
+    @Override
+    public boolean writeChar(byte[] data) {
+        for(MyBleManager manager: this.managers) {
+            if(manager.isReady()) {
+                manager.writeChar(data);
+                break;
+            }
+        }
+        return false;
+    }
+}

android/java/gl/entan/giouibind/MainActivity.java

@@ -0,0 +1,65 @@
+package gl.entan.giouibind;
+
+import org.gioui.GioView;
+
+import android.app.Activity;
+import android.app.ActivityManager;
+import android.content.Context;
+import android.content.res.Configuration;
+import android.graphics.BitmapFactory;
+import android.os.Build;
+import android.os.Bundle;
+import android.util.Log;
+import go.Seq;
+import gl.entan.giouibind.mobile.Mobile;
+
+public class MainActivity extends Activity {
+  private GioView view;
+
+	@Override public void onCreate(Bundle savedInstanceState) {
+		super.onCreate(savedInstanceState);
+		setContentView(R.layout.main);
+		Context context = getApplicationContext();
+		Seq.setContext(context);
+
+		this.view = findViewById(R.id.gioview);
+
+		if (Build.VERSION.SDK_INT >= Build.VERSION_CODES.LOLLIPOP) {
+			this.setTaskDescription(
+				new ActivityManager.TaskDescription(
+					null, // Leave the default title.
+					BitmapFactory.decodeResource(getResources(), R.drawable.icon)
+			));
+		}
+	}
+
+	@Override public void onDestroy() {
+		view.destroy();
+		super.onDestroy();
+	}
+
+	@Override public void onStart() {
+		super.onStart();
+		view.start();
+	}
+
+	@Override public void onStop() {
+		view.stop();
+		super.onStop();
+	}
+
+	@Override public void onConfigurationChanged(Configuration c) {
+		super.onConfigurationChanged(c);
+		view.configurationChanged();
+	}
+
+	@Override public void onLowMemory() {
+		super.onLowMemory();
+		GioView.onLowMemory();
+	}
+
+	@Override public void onBackPressed() {
+		if (!view.backPressed())
+			super.onBackPressed();
+	}
+}

android/java/gl/entan/giouibind/MyBleManager.java

@@ -0,0 +1,114 @@
+package gl.entan.giouibind;
+
+import java.util.List;
+import java.util.UUID;
+
+import android.bluetooth.BluetoothDevice;
+import android.bluetooth.BluetoothGatt;
+import android.bluetooth.BluetoothGattCallback;
+import android.bluetooth.BluetoothGattCharacteristic;
+import android.bluetooth.BluetoothGattService;
+import android.bluetooth.BluetoothProfile;
+import android.content.Context;
+import android.os.Handler;
+import android.util.Log;
+import gl.entan.giouibind.mobile.Mobile;
+import no.nordicsemi.android.ble.BleManager;
+
+public class MyBleManager extends BleManager {
+    public static UUID shortServiceUuid = UUID.fromString("FIXME");
+    public static UUID longServiceUuid = UUID.fromString("FIXME");
+
+    public static UUID writeUuid = UUID.fromString("FIXME");
+    public static UUID readUuid = UUID.fromString("FIXME");
+
+    private BluetoothGattService service;
+    private BluetoothGattCharacteristic writeChr;
+    private BluetoothGattCharacteristic readChr;
+
+    public MyBleManager(Context context) {
+        super(context);
+    }
+
+    public MyBleManager(Context context, Handler handler) {
+        super(context, handler);
+    }
+
+    public void readChar() {
+        readCharacteristic(readChr)
+            .with((d, data) -> {
+                Mobile.bluetoothGotData(data.getValue());
+            })
+            .enqueue();
+    }
+
+    public boolean writeChar(byte[] data) {
+        writeCharacteristic(writeChr, data, BluetoothGattCharacteristic.WRITE_TYPE_DEFAULT).enqueue();
+        return true;
+    }
+
+    @Override
+    protected BleManagerGattCallback getGattCallback() {
+        return new MyGattCallbackImpl();
+    }
+
+    @Override
+    public int getMinLogPriority() {
+        return Log.WARN;
+    }
+
+    @Override
+    public void log(int priority, String message) {
+        Log.println(priority, "bananas", message);
+    }
+
+    private class MyGattCallbackImpl extends BleManagerGattCallback {
+        @Override
+        protected void onCharacteristicNotified(BluetoothGatt gatt, BluetoothGattCharacteristic characteristic) {
+            Mobile.bluetoothGotData(characteristic.getValue());
+            super.onCharacteristicNotified(gatt, characteristic);
+        }
+
+        @Override
+        protected boolean isRequiredServiceSupported(BluetoothGatt gatt) {
+            BluetoothGattService s = gatt.getService(shortServiceUuid);
+            if(s == null) {
+                s = gatt.getService(longServiceUuid);
+            }
+            if(s == null) {
+                Log.i("bananas", "Could not find GATT service");
+                Mobile.finishedConnect(false, "");
+                return false;
+            }
+
+            service = s;
+            writeChr = s.getCharacteristic(writeUuid);
+            readChr = s.getCharacteristic(readUuid);
+
+            return service != null
+                && writeChr != null 
+                && readChr != null;
+        }
+
+        @Override
+        protected void initialize() {
+            requestMtu(165)
+                .done(d -> {
+                    enableNotifications(readChr).enqueue();
+
+                    writeCharacteristic(writeChr, new byte[] { 0x04, (byte)0x95, 0x06, 0x03 }, BluetoothGattCharacteristic.WRITE_TYPE_DEFAULT)
+                    .done(g -> {
+                        Mobile.finishedConnect(true, g.getName());
+                    })
+                    .enqueue();
+                })
+                .enqueue();
+        }
+
+        @Override
+        protected void onServicesInvalidated() {
+            readChr = writeChr = null;
+        }
+    }
+
+}

android/java/org/gioui/Gio.java

@@ -0,0 +1,71 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package org.gioui;
+
+import android.content.ClipboardManager;
+import android.content.ClipData;
+import android.content.Context;
+import android.os.Handler;
+import android.os.Looper;
+import gl.entan.giouibind.GodObject;
+import gl.entan.giouibind.mobile.Mobile;
+
+import java.io.UnsupportedEncodingException;
+
+public final class Gio {
+	private static final Object initLock = new Object();
+	private static boolean jniLoaded;
+	private static final Handler handler = new Handler(Looper.getMainLooper());
+
+	/**
+	 * init loads and initializes the Go native library and runs
+	 * the Go main function.
+	 *
+	 * It is exported for use by Android apps that need to run Go code
+	 * outside the lifecycle of the Gio activity.
+	 */
+	public static synchronized void init(Context appCtx) {
+		synchronized (initLock) {
+			if (jniLoaded) {
+				return;
+			}
+			String dataDir = appCtx.getFilesDir().getAbsolutePath();
+			byte[] dataDirUTF8;
+			try {
+				dataDirUTF8 = dataDir.getBytes("UTF-8");
+			} catch (UnsupportedEncodingException e) {
+				throw new RuntimeException(e);
+			}
+			System.loadLibrary("gojni");
+            Mobile.inventGod(GodObject.getGodObject(appCtx));
+			runGoMain(dataDirUTF8, appCtx);
+			jniLoaded = true;
+		}
+	}
+
+	static private native void runGoMain(byte[] dataDir, Context context);
+
+	static void writeClipboard(Context ctx, String s) {
+		ClipboardManager m = (ClipboardManager)ctx.getSystemService(Context.CLIPBOARD_SERVICE);
+		m.setPrimaryClip(ClipData.newPlainText(null, s));
+	}
+
+	static String readClipboard(Context ctx) {
+		ClipboardManager m = (ClipboardManager)ctx.getSystemService(Context.CLIPBOARD_SERVICE);
+		ClipData c = m.getPrimaryClip();
+		if (c == null || c.getItemCount() < 1) {
+			return null;
+		}
+		return c.getItemAt(0).coerceToText(ctx).toString();
+	}
+
+	static void wakeupMainThread() {
+		handler.post(new Runnable() {
+			@Override public void run() {
+				scheduleMainFuncs();
+			}
+		});
+	}
+
+	static private native void scheduleMainFuncs();
+}

android/java/org/gioui/GioView.java

@@ -0,0 +1,752 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package org.gioui;
+
+import android.app.Activity;
+import android.content.Context;
+import android.graphics.Color;
+import android.graphics.Matrix;
+import android.graphics.Rect;
+import android.os.Build;
+import android.os.Bundle;
+import android.os.Handler;
+import android.os.SystemClock;
+import android.text.TextUtils;
+import android.util.AttributeSet;
+import android.util.TypedValue;
+import android.view.Choreographer;
+import android.view.KeyEvent;
+import android.view.MotionEvent;
+import android.view.PointerIcon;
+import android.view.Surface;
+import android.view.SurfaceHolder;
+import android.view.SurfaceView;
+import android.view.View;
+import android.view.ViewConfiguration;
+import android.view.Window;
+import android.view.WindowInsetsController;
+import android.view.WindowManager;
+import android.view.accessibility.AccessibilityEvent;
+import android.view.accessibility.AccessibilityManager;
+import android.view.accessibility.AccessibilityNodeInfo;
+import android.view.accessibility.AccessibilityNodeProvider;
+import android.view.inputmethod.CompletionInfo;
+import android.view.inputmethod.CorrectionInfo;
+import android.view.inputmethod.CursorAnchorInfo;
+import android.view.inputmethod.EditorInfo;
+import android.view.inputmethod.ExtractedText;
+import android.view.inputmethod.ExtractedTextRequest;
+import android.view.inputmethod.InputConnection;
+import android.view.inputmethod.InputContentInfo;
+import android.view.inputmethod.InputMethodManager;
+import gl.entan.giouibind.GodObject;
+import gl.entan.giouibind.mobile.Mobile;
+
+public final class GioView extends SurfaceView implements Choreographer.FrameCallback {
+	private static boolean jniLoaded;
+
+	private final SurfaceHolder.Callback surfCallbacks;
+	private final View.OnFocusChangeListener focusCallback;
+	private final InputMethodManager imm;
+	private final float scrollXScale;
+	private final float scrollYScale;
+	private int keyboardHint;
+	private AccessibilityManager accessManager;
+
+	private long nhandle;
+
+	public GioView(Context context) {
+		this(context, null);
+	}
+
+	public GioView(Context context, AttributeSet attrs) {
+		super(context, attrs);
+		if (Build.VERSION.SDK_INT >= Build.VERSION_CODES.LOLLIPOP) {
+			setSystemUiVisibility(View.SYSTEM_UI_FLAG_LAYOUT_HIDE_NAVIGATION | View.SYSTEM_UI_FLAG_LAYOUT_STABLE);
+		}
+		setLayoutParams(new WindowManager.LayoutParams(WindowManager.LayoutParams.MATCH_PARENT, WindowManager.LayoutParams.MATCH_PARENT));
+
+		// Late initialization of the Go runtime to wait for a valid context.
+		Gio.init(context);
+
+		// Set background color to transparent to avoid a flickering
+		// issue on ChromeOS.
+		setBackgroundColor(Color.argb(0, 0, 0, 0));
+
+		ViewConfiguration conf = ViewConfiguration.get(context);
+		if (Build.VERSION.SDK_INT >= Build.VERSION_CODES.O) {
+			scrollXScale = conf.getScaledHorizontalScrollFactor();
+			scrollYScale = conf.getScaledVerticalScrollFactor();
+
+			// The platform focus highlight is not aware of Gio's widgets.
+			setDefaultFocusHighlightEnabled(false);
+		} else {
+			float listItemHeight = 48; // dp
+			float px = TypedValue.applyDimension(
+				TypedValue.COMPLEX_UNIT_DIP,
+				listItemHeight,
+				getResources().getDisplayMetrics()
+			);
+			scrollXScale = px;
+			scrollYScale = px;
+		}
+
+		accessManager = (AccessibilityManager)context.getSystemService(Context.ACCESSIBILITY_SERVICE);
+		imm = (InputMethodManager)context.getSystemService(Context.INPUT_METHOD_SERVICE);
+		nhandle = onCreateView(this);
+		setFocusable(true);
+		setFocusableInTouchMode(true);
+		focusCallback = new View.OnFocusChangeListener() {
+			@Override public void onFocusChange(View v, boolean focus) {
+				GioView.this.onFocusChange(nhandle, focus);
+			}
+		};
+		setOnFocusChangeListener(focusCallback);
+		surfCallbacks = new SurfaceHolder.Callback() {
+			@Override public void surfaceCreated(SurfaceHolder holder) {
+				// Ignore; surfaceChanged is guaranteed to be called immediately after this.
+			}
+			@Override public void surfaceChanged(SurfaceHolder holder, int format, int width, int height) {
+				onSurfaceChanged(nhandle, getHolder().getSurface());
+			}
+			@Override public void surfaceDestroyed(SurfaceHolder holder) {
+				onSurfaceDestroyed(nhandle);
+			}
+		};
+		getHolder().addCallback(surfCallbacks);
+	}
+
+	@Override public boolean onKeyDown(int keyCode, KeyEvent event) {
+		if (nhandle != 0) {
+			onKeyEvent(nhandle, keyCode, event.getUnicodeChar(), true, event.getEventTime());
+		}
+		return false;
+	}
+
+	@Override public boolean onKeyUp(int keyCode, KeyEvent event) {
+		if (nhandle != 0) {
+			onKeyEvent(nhandle, keyCode, event.getUnicodeChar(), false, event.getEventTime());
+		}
+		return false;
+	}
+
+	@Override public boolean onGenericMotionEvent(MotionEvent event) {
+		dispatchMotionEvent(event);
+		return true;
+	}
+
+	@Override public boolean onTouchEvent(MotionEvent event) {
+		// Ask for unbuffered events. Flutter and Chrome do it
+		// so assume it's good for us as well.
+		if (Build.VERSION.SDK_INT >= Build.VERSION_CODES.LOLLIPOP) {
+			requestUnbufferedDispatch(event);
+		}
+
+		dispatchMotionEvent(event);
+		return true;
+	}
+
+	private void setCursor(int id) {
+		if (Build.VERSION.SDK_INT < Build.VERSION_CODES.N) {
+			return;
+		}
+		PointerIcon pointerIcon = PointerIcon.getSystemIcon(getContext(), id);
+		setPointerIcon(pointerIcon);
+	}
+
+	private void setOrientation(int id, int fallback) {
+		if (Build.VERSION.SDK_INT < Build.VERSION_CODES.JELLY_BEAN_MR2) {
+			id = fallback;
+		}
+		((Activity) this.getContext()).setRequestedOrientation(id);
+	}
+
+	private void setFullscreen(boolean enabled) {
+		int flags = this.getSystemUiVisibility();
+		if (enabled) {
+		   flags |= SYSTEM_UI_FLAG_IMMERSIVE_STICKY;
+		   flags |= SYSTEM_UI_FLAG_HIDE_NAVIGATION;
+		   flags |= SYSTEM_UI_FLAG_FULLSCREEN;
+		   flags |= SYSTEM_UI_FLAG_LAYOUT_FULLSCREEN;
+		} else {
+		   flags &= ~SYSTEM_UI_FLAG_IMMERSIVE_STICKY;
+		   flags &= ~SYSTEM_UI_FLAG_HIDE_NAVIGATION;
+		   flags &= ~SYSTEM_UI_FLAG_FULLSCREEN;
+		   flags &= ~SYSTEM_UI_FLAG_LAYOUT_FULLSCREEN;
+		}
+		this.setSystemUiVisibility(flags);
+	}
+
+	private enum Bar {
+		NAVIGATION,
+		STATUS,
+	}
+
+	private void setBarColor(Bar t, int color, int luminance) {
+		if (Build.VERSION.SDK_INT < Build.VERSION_CODES.LOLLIPOP) {
+			return;
+		}
+
+		Window window = ((Activity) this.getContext()).getWindow();
+
+		int insetsMask;
+		int viewMask;
+
+		switch (t) {
+		case STATUS:
+			insetsMask = WindowInsetsController.APPEARANCE_LIGHT_STATUS_BARS;
+			viewMask = View.SYSTEM_UI_FLAG_LIGHT_STATUS_BAR;
+			window.setStatusBarColor(color);
+			break;
+		case NAVIGATION:
+			insetsMask = WindowInsetsController.APPEARANCE_LIGHT_NAVIGATION_BARS;
+			viewMask = View.SYSTEM_UI_FLAG_LIGHT_NAVIGATION_BAR;
+			window.setNavigationBarColor(color);
+			break;
+		default:
+			throw new RuntimeException("invalid bar type");
+		}
+
+		if (Build.VERSION.SDK_INT < Build.VERSION_CODES.M) {
+			return;
+		}
+
+		if (Build.VERSION.SDK_INT < Build.VERSION_CODES.R) {
+			int flags = this.getSystemUiVisibility();
+			if (luminance > 128) {
+			   flags |=  viewMask;
+			} else {
+			   flags &= ~viewMask;
+			}
+			this.setSystemUiVisibility(flags);
+			return;
+		}
+
+		WindowInsetsController insetsController = window.getInsetsController();
+		if (insetsController == null) {
+			return;
+		}
+		if (luminance > 128) {
+			insetsController.setSystemBarsAppearance(insetsMask, insetsMask);
+		} else {
+			insetsController.setSystemBarsAppearance(0, insetsMask);
+		}
+	}
+
+	private void setStatusColor(int color, int luminance) {
+		this.setBarColor(Bar.STATUS, color, luminance);
+	}
+
+	private void setNavigationColor(int color, int luminance) {
+		this.setBarColor(Bar.NAVIGATION, color, luminance);
+	}
+
+	@Override protected boolean dispatchHoverEvent(MotionEvent event) {
+		if (!accessManager.isTouchExplorationEnabled()) {
+			return super.dispatchHoverEvent(event);
+		}
+		switch (event.getAction()) {
+		case MotionEvent.ACTION_HOVER_ENTER:
+			// Fall through.
+		case MotionEvent.ACTION_HOVER_MOVE:
+			onTouchExploration(nhandle, event.getX(), event.getY());
+			break;
+		case MotionEvent.ACTION_HOVER_EXIT:
+			onExitTouchExploration(nhandle);
+			break;
+		}
+		return true;
+	}
+
+	void sendA11yEvent(int eventType, int viewId) {
+		if (!accessManager.isEnabled()) {
+			return;
+		}
+		AccessibilityEvent event = obtainA11yEvent(eventType, viewId);
+		getParent().requestSendAccessibilityEvent(this, event);
+	}
+
+	AccessibilityEvent obtainA11yEvent(int eventType, int viewId) {
+		AccessibilityEvent event = AccessibilityEvent.obtain(eventType);
+		event.setPackageName(getContext().getPackageName());
+		event.setSource(this, viewId);
+		return event;
+	}
+
+	boolean isA11yActive() {
+		return accessManager.isEnabled();
+	}
+
+	void sendA11yChange(int viewId) {
+		if (!accessManager.isEnabled()) {
+			return;
+		}
+		AccessibilityEvent event = obtainA11yEvent(AccessibilityEvent.TYPE_WINDOW_CONTENT_CHANGED, viewId);
+		if (Build.VERSION.SDK_INT >= Build.VERSION_CODES.KITKAT) {
+			event.setContentChangeTypes(AccessibilityEvent.CONTENT_CHANGE_TYPE_SUBTREE);
+		}
+		getParent().requestSendAccessibilityEvent(this, event);
+	}
+
+	private void dispatchMotionEvent(MotionEvent event) {
+		if (nhandle == 0) {
+			return;
+		}
+		for (int j = 0; j < event.getHistorySize(); j++) {
+			long time = event.getHistoricalEventTime(j);
+			for (int i = 0; i < event.getPointerCount(); i++) {
+				onTouchEvent(
+						nhandle,
+						event.ACTION_MOVE,
+						event.getPointerId(i),
+						event.getToolType(i),
+						event.getHistoricalX(i, j),
+						event.getHistoricalY(i, j),
+						scrollXScale*event.getHistoricalAxisValue(MotionEvent.AXIS_HSCROLL, i, j),
+						scrollYScale*event.getHistoricalAxisValue(MotionEvent.AXIS_VSCROLL, i, j),
+						event.getButtonState(),
+						time);
+			}
+		}
+		int act = event.getActionMasked();
+		int idx = event.getActionIndex();
+		for (int i = 0; i < event.getPointerCount(); i++) {
+			int pact = event.ACTION_MOVE;
+			if (i == idx) {
+				pact = act;
+			}
+			onTouchEvent(
+					nhandle,
+					pact,
+					event.getPointerId(i),
+					event.getToolType(i),
+					event.getX(i), event.getY(i),
+					scrollXScale*event.getAxisValue(MotionEvent.AXIS_HSCROLL, i),
+					scrollYScale*event.getAxisValue(MotionEvent.AXIS_VSCROLL, i),
+					event.getButtonState(),
+					event.getEventTime());
+		}
+	}
+
+	@Override public InputConnection onCreateInputConnection(EditorInfo editor) {
+		Snippet snip = getSnippet();
+		editor.inputType = this.keyboardHint;
+		editor.imeOptions = EditorInfo.IME_FLAG_NO_FULLSCREEN | EditorInfo.IME_FLAG_NO_EXTRACT_UI;
+		editor.initialSelStart = imeToUTF16(nhandle, imeSelectionStart(nhandle));
+		editor.initialSelEnd = imeToUTF16(nhandle, imeSelectionEnd(nhandle));
+		int selStart = editor.initialSelStart - snip.offset;
+		editor.initialCapsMode = TextUtils.getCapsMode(snip.snippet, selStart, this.keyboardHint);
+		imeSetComposingRegion(nhandle, -1, -1);
+		return new GioInputConnection();
+	}
+
+	void setInputHint(int hint) {
+		if (hint == this.keyboardHint) {
+			return;
+		}
+		this.keyboardHint = hint;
+		restartInput();
+	}
+
+	void showTextInput() {
+		GioView.this.requestFocus();
+		imm.showSoftInput(GioView.this, 0);
+	}
+
+	void hideTextInput() {
+		imm.hideSoftInputFromWindow(getWindowToken(), 0);
+	}
+
+	@Override protected boolean fitSystemWindows(Rect insets) {
+		if (nhandle != 0) {
+			onWindowInsets(nhandle, insets.top, insets.right, insets.bottom, insets.left);
+		}
+		return true;
+	}
+
+	void postFrameCallback() {
+		Choreographer.getInstance().removeFrameCallback(this);
+		Choreographer.getInstance().postFrameCallback(this);
+	}
+
+	@Override public void doFrame(long nanos) {
+		if (nhandle != 0) {
+			onFrameCallback(nhandle);
+		}
+	}
+
+	int getDensity() {
+		return getResources().getDisplayMetrics().densityDpi;
+	}
+
+	float getFontScale() {
+		return getResources().getConfiguration().fontScale;
+	}
+
+	public void start() {
+		if (nhandle != 0) {
+			onStartView(nhandle);
+		}
+	}
+
+	public void stop() {
+		if (nhandle != 0) {
+			onStopView(nhandle);
+		}
+	}
+
+	public void destroy() {
+		if (nhandle != 0) {
+			onDestroyView(nhandle);
+		}
+	}
+
+	protected void unregister() {
+		setOnFocusChangeListener(null);
+		getHolder().removeCallback(surfCallbacks);
+		nhandle = 0;
+	}
+
+	public void configurationChanged() {
+		if (nhandle != 0) {
+			onConfigurationChanged(nhandle);
+		}
+	}
+
+	public boolean backPressed() {
+		if (nhandle == 0) {
+			return false;
+		}
+		return onBack(nhandle);
+	}
+
+	void restartInput() {
+		imm.restartInput(this);
+	}
+
+	void updateSelection() {
+		int selStart = imeToUTF16(nhandle, imeSelectionStart(nhandle));
+		int selEnd = imeToUTF16(nhandle, imeSelectionEnd(nhandle));
+		int compStart = imeToUTF16(nhandle, imeComposingStart(nhandle));
+		int compEnd = imeToUTF16(nhandle, imeComposingEnd(nhandle));
+		imm.updateSelection(this, selStart, selEnd, compStart, compEnd);
+	}
+
+	void updateCaret(float m00, float m01, float m02, float m10, float m11, float m12, float caretX, float caretTop, float caretBase, float caretBottom) {
+		if (Build.VERSION.SDK_INT < Build.VERSION_CODES.LOLLIPOP) {
+			return;
+		}
+		Matrix m = new Matrix();
+		m.setValues(new float[]{m00, m01, m02, m10, m11, m12, 0.0f, 0.0f, 1.0f});
+		m.setConcat(getMatrix(), m);
+		int selStart = imeSelectionStart(nhandle);
+		int selEnd = imeSelectionEnd(nhandle);
+		int compStart = imeComposingStart(nhandle);
+		int compEnd = imeComposingEnd(nhandle);
+		Snippet snip = getSnippet();
+		String composing = "";
+		if (compStart != -1) {
+			composing = snip.substringRunes(compStart, compEnd);
+		}
+		CursorAnchorInfo inf = new CursorAnchorInfo.Builder()
+			.setMatrix(m)
+			.setComposingText(imeToUTF16(nhandle, compStart), composing)
+			.setSelectionRange(imeToUTF16(nhandle, selStart), imeToUTF16(nhandle, selEnd))
+			.setInsertionMarkerLocation(caretX, caretTop, caretBase, caretBottom, 0)
+			.build();
+		imm.updateCursorAnchorInfo(this, inf);
+	}
+
+	static private native long onCreateView(GioView view);
+	static private native void onDestroyView(long handle);
+	static private native void onStartView(long handle);
+	static private native void onStopView(long handle);
+	static private native void onSurfaceDestroyed(long handle);
+	static private native void onSurfaceChanged(long handle, Surface surface);
+	static private native void onConfigurationChanged(long handle);
+	static private native void onWindowInsets(long handle, int top, int right, int bottom, int left);
+	static public native void onLowMemory();
+	static private native void onTouchEvent(long handle, int action, int pointerID, int tool, float x, float y, float scrollX, float scrollY, int buttons, long time);
+	static private native void onKeyEvent(long handle, int code, int character, boolean pressed, long time);
+	static private native void onFrameCallback(long handle);
+	static private native boolean onBack(long handle);
+	static private native void onFocusChange(long handle, boolean focus);
+	static private native AccessibilityNodeInfo initializeAccessibilityNodeInfo(long handle, int viewId, int screenX, int screenY, AccessibilityNodeInfo info);
+	static private native void onTouchExploration(long handle, float x, float y);
+	static private native void onExitTouchExploration(long handle);
+	static private native void onA11yFocus(long handle, int viewId);
+	static private native void onClearA11yFocus(long handle, int viewId);
+	static private native void imeSetSnippet(long handle, int start, int end);
+	static private native String imeSnippet(long handle);
+	static private native int imeSnippetStart(long handle);
+	static private native int imeSelectionStart(long handle);
+	static private native int imeSelectionEnd(long handle);
+	static private native int imeComposingStart(long handle);
+	static private native int imeComposingEnd(long handle);
+	static private native int imeReplace(long handle, int start, int end, String text);
+	static private native int imeSetSelection(long handle, int start, int end);
+	static private native int imeSetComposingRegion(long handle, int start, int end);
+	// imeToRunes converts the Java character index into runes (Java code points).
+	static private native int imeToRunes(long handle, int chars);
+	// imeToUTF16 converts the rune index into Java characters.
+	static private native int imeToUTF16(long handle, int runes);
+
+	private class GioInputConnection implements InputConnection {
+		private int batchDepth;
+
+		@Override public boolean beginBatchEdit() {
+			batchDepth++;
+			return true;
+		}
+
+		@Override public boolean endBatchEdit() {
+			batchDepth--;
+			return batchDepth > 0;
+		}
+
+		@Override public boolean clearMetaKeyStates(int states) {
+			return false;
+		}
+
+		@Override public boolean commitCompletion(CompletionInfo text) {
+			return false;
+		}
+
+		@Override public boolean commitCorrection(CorrectionInfo info) {
+			return false;
+		}
+
+		@Override public boolean commitText(CharSequence text, int cursor) {
+			setComposingText(text, cursor);
+			return finishComposingText();
+		}
+
+		@Override public boolean deleteSurroundingText(int beforeChars, int afterChars) {
+			// translate before and after to runes.
+			int selStart = imeSelectionStart(nhandle);
+			int selEnd = imeSelectionEnd(nhandle);
+			int before = selStart - imeToRunes(nhandle, imeToUTF16(nhandle, selStart) - beforeChars);
+			int after = selEnd - imeToRunes(nhandle, imeToUTF16(nhandle, selEnd) - afterChars);
+			return deleteSurroundingTextInCodePoints(before, after);
+		}
+
+		@Override public boolean finishComposingText() {
+			imeSetComposingRegion(nhandle, -1, -1);
+			return true;
+		}
+
+		@Override public int getCursorCapsMode(int reqModes) {
+			Snippet snip = getSnippet();
+			int selStart = imeSelectionStart(nhandle);
+			return TextUtils.getCapsMode(snip.snippet, imeToUTF16(nhandle, selStart), reqModes);
+		}
+
+		@Override public ExtractedText getExtractedText(ExtractedTextRequest request, int flags) {
+			return null;
+		}
+
+		@Override public CharSequence getSelectedText(int flags) {
+			Snippet snip = getSnippet();
+			int selStart = imeSelectionStart(nhandle);
+			int selEnd = imeSelectionEnd(nhandle);
+			String sub = snip.substringRunes(selStart, selEnd);
+			return sub;
+		}
+
+		@Override public CharSequence getTextAfterCursor(int n, int flags) {
+			Snippet snip = getSnippet();
+			int selStart = imeSelectionStart(nhandle);
+			int selEnd = imeSelectionEnd(nhandle);
+			// n are in Java characters, but in worst case we'll just ask for more runes
+			// than wanted.
+			imeSetSnippet(nhandle, selStart - n, selEnd + n);
+			int start = selEnd;
+			int end = imeToRunes(nhandle, imeToUTF16(nhandle, selEnd) + n);
+			String ret = snip.substringRunes(start, end);
+			return ret;
+		}
+
+		@Override public CharSequence getTextBeforeCursor(int n, int flags) {
+			Snippet snip = getSnippet();
+			int selStart = imeSelectionStart(nhandle);
+			int selEnd = imeSelectionEnd(nhandle);
+			// n are in Java characters, but in worst case we'll just ask for more runes
+			// than wanted.
+			imeSetSnippet(nhandle, selStart - n, selEnd + n);
+			int start = imeToRunes(nhandle, imeToUTF16(nhandle, selStart) - n);
+			int end = selStart;
+			String ret = snip.substringRunes(start, end);
+			return ret;
+		}
+
+		@Override public boolean performContextMenuAction(int id) {
+			return false;
+		}
+
+		@Override public boolean performEditorAction(int editorAction) {
+			long eventTime = SystemClock.uptimeMillis();
+			// Translate to enter key.
+			onKeyEvent(nhandle, KeyEvent.KEYCODE_ENTER, '\n', true, eventTime);
+			onKeyEvent(nhandle, KeyEvent.KEYCODE_ENTER, '\n', false, eventTime);
+			return true;
+		}
+
+		@Override public boolean performPrivateCommand(String action, Bundle data) {
+			return false;
+		}
+
+		@Override public boolean reportFullscreenMode(boolean enabled) {
+			return false;
+		}
+
+		@Override public boolean sendKeyEvent(KeyEvent event) {
+			boolean pressed = event.getAction() == KeyEvent.ACTION_DOWN;
+			onKeyEvent(nhandle, event.getKeyCode(), event.getUnicodeChar(), pressed, event.getEventTime());
+			return true;
+		}
+
+		@Override public boolean setComposingRegion(int startChars, int endChars) {
+			int compStart = imeToRunes(nhandle, startChars);
+			int compEnd = imeToRunes(nhandle, endChars);
+			imeSetComposingRegion(nhandle, compStart, compEnd);
+			return true;
+		}
+
+		@Override public boolean setComposingText(CharSequence text, int relCursor) {
+			int start = imeComposingStart(nhandle);
+			int end = imeComposingEnd(nhandle);
+			if (start == -1 || end == -1) {
+				start = imeSelectionStart(nhandle);
+				end = imeSelectionEnd(nhandle);
+			}
+			String str = text.toString();
+			imeReplace(nhandle, start, end, str);
+			int cursor = start;
+			int runes = str.codePointCount(0, str.length());
+			if (relCursor > 0) {
+				cursor += runes;
+				relCursor--;
+			}
+			imeSetComposingRegion(nhandle, start, start + runes);
+
+			// Move cursor.
+			Snippet snip = getSnippet();
+			cursor = imeToRunes(nhandle, imeToUTF16(nhandle, cursor) + relCursor);
+			imeSetSelection(nhandle, cursor, cursor);
+			return true;
+		}
+
+		@Override public boolean setSelection(int startChars, int endChars) {
+			int start = imeToRunes(nhandle, startChars);
+			int end = imeToRunes(nhandle, endChars);
+			imeSetSelection(nhandle, start, end);
+			return true;
+		}
+
+		/*@Override*/ public boolean requestCursorUpdates(int cursorUpdateMode) {
+			// We always provide cursor updates.
+			return true;
+		}
+
+		/*@Override*/ public void closeConnection() {
+		}
+
+		/*@Override*/ public Handler getHandler() {
+			return null;
+		}
+
+		/*@Override*/ public boolean commitContent(InputContentInfo info, int flags, Bundle opts) {
+			return false;
+		}
+
+		/*@Override*/ public boolean deleteSurroundingTextInCodePoints(int before, int after) {
+			if (after > 0) {
+				int selEnd = imeSelectionEnd(nhandle);
+				imeReplace(nhandle, selEnd, selEnd + after, "");
+			}
+			if (before > 0) {
+				int selStart = imeSelectionStart(nhandle);
+				imeReplace(nhandle, selStart - before, selStart, "");
+			}
+			return true;
+		}
+	}
+
+	private Snippet getSnippet() {
+		Snippet snip = new Snippet();
+		snip.snippet = imeSnippet(nhandle);
+		snip.offset = imeSnippetStart(nhandle);
+		return snip;
+	}
+
+	// Snippet is like android.view.inputmethod.SurroundingText but available for Android < 31.
+	private static class Snippet {
+		String snippet;
+		// offset of snippet into the entire editor content. It is in runes because we won't require
+		// Gio editors to keep track of UTF-16 offsets. The distinction won't matter in practice because IMEs only
+		// ever see snippets.
+		int offset;
+
+		// substringRunes returns the substring from start to end in runes. The resuls is
+		// truncated to the snippet.
+		String substringRunes(int start, int end) {
+			start -= this.offset;
+			end -= this.offset;
+			int runes = snippet.codePointCount(0, snippet.length());
+			if (start < 0) {
+				start = 0;
+			}
+			if (end < 0) {
+				end = 0;
+			}
+			if (start > runes) {
+				start = runes;
+			}
+			if (end > runes) {
+				end = runes;
+			}
+			return snippet.substring(
+				snippet.offsetByCodePoints(0, start),
+				snippet.offsetByCodePoints(0, end)
+			);
+		}
+	}
+
+	@Override public AccessibilityNodeProvider getAccessibilityNodeProvider() {
+		return new AccessibilityNodeProvider() {
+			private final int[] screenOff = new int[2];
+
+			@Override public AccessibilityNodeInfo createAccessibilityNodeInfo(int viewId) {
+				AccessibilityNodeInfo info = null;
+				if (viewId == View.NO_ID) {
+					info = AccessibilityNodeInfo.obtain(GioView.this);
+					GioView.this.onInitializeAccessibilityNodeInfo(info);
+				} else {
+					info = AccessibilityNodeInfo.obtain(GioView.this, viewId);
+					info.setPackageName(getContext().getPackageName());
+					info.setVisibleToUser(true);
+				}
+				GioView.this.getLocationOnScreen(screenOff);
+				info = GioView.this.initializeAccessibilityNodeInfo(nhandle, viewId, screenOff[0], screenOff[1], info);
+				return info;
+			}
+
+			@Override public boolean performAction(int viewId, int action, Bundle arguments) {
+				if (viewId == View.NO_ID) {
+					return GioView.this.performAccessibilityAction(action, arguments);
+				}
+				switch (action) {
+				case AccessibilityNodeInfo.ACTION_ACCESSIBILITY_FOCUS:
+					GioView.this.onA11yFocus(nhandle, viewId);
+					GioView.this.sendA11yEvent(AccessibilityEvent.TYPE_VIEW_ACCESSIBILITY_FOCUSED, viewId);
+					return true;
+				case AccessibilityNodeInfo.ACTION_CLEAR_ACCESSIBILITY_FOCUS:
+					GioView.this.onClearA11yFocus(nhandle, viewId);
+					GioView.this.sendA11yEvent(AccessibilityEvent.TYPE_VIEW_ACCESSIBILITY_FOCUS_CLEARED, viewId);
+					return true;
+				}
+				return false;
+			}
+		};
+	}
+}

android/res/layout/main.xml

@@ -0,0 +1,11 @@
+<RelativeLayout xmlns:android="http://schemas.android.com/apk/res/android"
+    xmlns:tools="http://schemas.android.com/tools"
+    android:layout_width="match_parent"
+    android:layout_height="match_parent"
+    tools:context=".MainActivity"
+    android:keepScreenOn="true">
+   <org.gioui.GioView
+           android:id="@+id/gioview"
+           android:layout_width="match_parent"
+           android:layout_height="match_parent" />
+</RelativeLayout>

android/res/values-de/strings.xml

@@ -0,0 +1,5 @@
+<?xml version="1.0" encoding="utf-8"?>
+<resources>
+    <string name="channel_name">Benachrichtigungen</string>
+    <string name="channel_description">Wunder</string>
+</resources>

android/res/values/strings.xml

@@ -0,0 +1,5 @@
+<?xml version="1.0" encoding="utf-8"?>
+<resources>
+    <string name="channel_name">Notifications</string>
+    <string name="channel_description">Wow</string>
+</resources>

app_state.go

@@ -0,0 +1,21 @@
+package giouibind
+
+import (
+	"gioui.org/app"
+	"github.com/empathicqubit/giouibind/native"
+)
+
+type AppSettings struct {
+}
+
+type AppState struct {
+	settings         *AppSettings
+	window           *app.Window
+	bluetoothEnabled bool
+	loaded           bool
+	connected        bool
+	connecting       bool
+	initted          bool
+	deviceName       string
+	nativeBridge     native.INativeBridge
+}

desktop/god.go

@@ -0,0 +1,51 @@
+package main
+
+import (
+	"strconv"
+	"strings"
+
+	"github.com/JuulLabs-OSS/cbgo"
+	"github.com/empathicqubit/giouibind/native"
+)
+
+var _ native.INativeBridge = (*NativeBridge)(nil)
+
+type NativeBridge struct {
+	fd       int
+	central  *cbgo.CentralManager
+	delegate *MyDelegate
+}
+
+// str2ba converts MAC address string representation to little-endian byte array
+func str2ba(addr string) [6]byte {
+	a := strings.Split(addr, ":")
+	var b [6]byte
+	for i, tmp := range a {
+		u, _ := strconv.ParseUint(tmp, 16, 8)
+		b[len(b)-1-i] = byte(u)
+	}
+	return b
+}
+
+func (god *NativeBridge) EnableBluetooth() bool {
+	central := cbgo.NewCentralManager(&cbgo.ManagerOpts{})
+	god.central = &central
+	delegate := &MyDelegate{}
+	god.delegate = delegate
+	god.central.SetDelegate(delegate)
+
+	return true
+}
+
+var localName string = ""
+
+func (god *NativeBridge) WriteChar(data []byte) bool {
+	if data == nil {
+		return false
+	}
+
+	return god.delegate.writeChar(data)
+}
+
+func (god *NativeBridge) ConnectToDevice() {
+}

desktop/main.go

@@ -0,0 +1,39 @@
+package main
+
+import (
+	"log"
+	"os"
+
+	"gioui.org/app"
+	"gioui.org/unit"
+	"github.com/empathicqubit/giouibind"
+	"github.com/empathicqubit/giouibind/native"
+)
+
+var nativeBridge *NativeBridge = &NativeBridge{}
+var dumbApp *giouibind.AppState = &giouibind.AppState{}
+
+func main() {
+	var igod native.INativeBridge = nativeBridge
+	log.Println("Called main")
+	err := dumbApp.Load()
+	if err != nil {
+		panic(err)
+	}
+	err = dumbApp.InventGod(igod)
+	if err != nil {
+		panic(err)
+	}
+
+	go func() {
+		w := app.NewWindow(
+			app.Size(unit.Dp(450), unit.Dp(800)),
+		)
+		err := dumbApp.RunApp(w)
+		if err != nil {
+			log.Fatal(err)
+		}
+		os.Exit(0)
+	}()
+	app.Main()
+}

desktop/my_delegate.go

@@ -0,0 +1,163 @@
+package main
+
+import (
+	"log"
+
+	"github.com/JuulLabs-OSS/cbgo"
+)
+
+type MyDelegate struct {
+	cbgo.CentralManagerDelegateBase
+	cbgo.PeripheralDelegateBase
+	connectedPeripheral *cbgo.Peripheral
+	service             *cbgo.Service
+	readChr             cbgo.Characteristic
+	writeChr            cbgo.Characteristic
+	encryptedChr        cbgo.Characteristic
+	pairingChr          cbgo.Characteristic
+	peripheral          *cbgo.Peripheral
+	finishedConnect     bool
+}
+
+var shortServiceUuid, _ = cbgo.ParseUUID16("FIXME")
+var longServiceUuid, _ = cbgo.ParseUUID("FIXME")
+
+var writeUuid, _ = cbgo.ParseUUID("FIXME")
+var readUuid, _ = cbgo.ParseUUID("FIXME")
+
+func (d *MyDelegate) writeChar(data []byte) bool {
+	if d.connectedPeripheral == nil {
+		return false
+	}
+
+	d.connectedPeripheral.WriteCharacteristic(data, d.writeChr, true)
+
+	return true
+}
+
+func (d *MyDelegate) readChar() []byte {
+	if d.connectedPeripheral == nil {
+		return []byte{}
+	}
+
+	d.connectedPeripheral.ReadCharacteristic(d.readChr)
+
+	return d.readChr.Value()
+}
+
+func (d *MyDelegate) CentralManagerDidUpdateState(cmgr cbgo.CentralManager) {
+	if cmgr.State() == cbgo.ManagerStatePoweredOn {
+		log.Println("Start scanning")
+		go func() {
+			nativeBridge.central.Scan([]cbgo.UUID{shortServiceUuid, longServiceUuid}, &cbgo.CentralManagerScanOpts{
+				AllowDuplicates:       false,
+				SolicitedServiceUUIDs: []cbgo.UUID{longServiceUuid, shortServiceUuid},
+			})
+		}()
+	}
+}
+
+func (d *MyDelegate) DidDiscoverPeripheral(cm cbgo.CentralManager, prph cbgo.Peripheral,
+	advFields cbgo.AdvFields, rssi int) {
+	log.Println("Found peripheral", prph.Name())
+	nativeBridge.central.Connect(prph, nil)
+}
+
+func (d *MyDelegate) DidConnectPeripheral(cm cbgo.CentralManager, prph cbgo.Peripheral) {
+	if d.connectedPeripheral == nil {
+		prph.SetDelegate(d)
+		prph.DiscoverServices([]cbgo.UUID{longServiceUuid, shortServiceUuid})
+	}
+}
+
+func (d *MyDelegate) DidFailToConnectPeripheral(cm cbgo.CentralManager, prph cbgo.Peripheral, err error) {
+	log.Printf("failed to connect: %v", err)
+}
+
+func (d *MyDelegate) DidDisconnectPeripheral(cm cbgo.CentralManager, prph cbgo.Peripheral, err error) {
+	log.Printf("peripheral disconnected: %v", err)
+}
+
+func (d *MyDelegate) DidDiscoverServices(prph cbgo.Peripheral, err error) {
+	if err != nil || len(prph.Services()) == 0 {
+		log.Println("Error discovering services", err)
+		dumbApp.FinishedConnect(false, "")
+		return
+	}
+
+	d.service = &prph.Services()[0]
+	prph.DiscoverCharacteristics([]cbgo.UUID{writeUuid, readUuid, encryptedUuid, pairingUuid}, *d.service)
+}
+
+func (d *MyDelegate) DidDiscoverCharacteristics(prph cbgo.Peripheral, svc cbgo.Service, err error) {
+	if svc.UUID().String() != d.service.UUID().String() {
+		return
+	}
+
+	if err != nil || len(svc.Characteristics()) == 0 {
+		log.Println("No characteristics")
+		dumbApp.FinishedConnect(false, "")
+		return
+	}
+
+	foundCount := 0
+	for _, chr := range d.service.Characteristics() {
+		log.Println("Found char", chr.UUID().String())
+
+		uuid := chr.UUID().String()
+		if uuid == writeUuid.String() {
+			d.writeChr = chr
+			foundCount += 1
+		} else if uuid == readUuid.String() {
+			d.readChr = chr
+			foundCount += 1
+                }
+	}
+
+	if foundCount < 2 || d.service == nil {
+		log.Println("Couldn't find matching service or characteristics")
+		dumbApp.FinishedConnect(false, "")
+		return
+	}
+
+	nativeBridge.central.StopScan()
+	d.connectedPeripheral = &prph
+
+	log.Println("Finished connect. Writing test command")
+
+	d.connectedPeripheral.SetNotify(true, d.readChr)
+}
+
+func (d *MyDelegate) DidDiscoverDescriptors(prph cbgo.Peripheral, chr cbgo.Characteristic, err error) {
+}
+
+func (d *MyDelegate) DidUpdateValueForCharacteristic(prph cbgo.Peripheral, chr cbgo.Characteristic, err error) {
+	if err != nil {
+		log.Printf("Error getting descriptor value: %s", chr.UUID().String())
+		return
+	}
+
+	if chr.Value() == nil {
+		return
+	}
+	data := chr.Value()
+
+	if chr.UUID().String() != d.readChr.UUID().String() {
+		return
+	}
+
+	dumbApp.BluetoothGotData(data)
+
+	if !d.finishedConnect {
+		d.finishedConnect = true
+		dumbApp.FinishedConnect(true, d.connectedPeripheral.Name())
+	}
+}
+
+func (d *MyDelegate) DidUpdateValueForDescriptor(prph cbgo.Peripheral, dsc cbgo.Descriptor, err error) {
+	if dsc.Characteristic().Service().UUID().String() != d.service.UUID().String() {
+		return
+	}
+
+	d.DidUpdateValueForCharacteristic(prph, dsc.Characteristic(), err)
+}

gio/cpu/abi.h

@@ -0,0 +1,91 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#define ALIGN(bytes, type) type __attribute__((aligned(bytes)))
+
+typedef ALIGN(8, uint8_t) byte8[8];
+typedef ALIGN(8, uint16_t) word4[4];
+typedef ALIGN(4, uint32_t) dword;
+typedef ALIGN(16, uint32_t) dword4[4];
+typedef ALIGN(8, uint64_t) qword;
+typedef ALIGN(16, uint64_t) qword2[2];
+typedef ALIGN(16, unsigned int) uint4[4];
+typedef ALIGN(8, uint32_t) dword2[2];
+typedef ALIGN(8, unsigned short) ushort4[4];
+typedef ALIGN(16, float) float4[4];
+typedef ALIGN(16, int) int4[4];
+
+typedef unsigned short half;
+
+typedef unsigned char bool;
+
+enum {
+	MAX_BOUND_DESCRIPTOR_SETS = 4,
+	MAX_DESCRIPTOR_SET_UNIFORM_BUFFERS_DYNAMIC = 8,
+	MAX_DESCRIPTOR_SET_STORAGE_BUFFERS_DYNAMIC = 4,
+	MAX_DESCRIPTOR_SET_COMBINED_BUFFERS_DYNAMIC =
+		MAX_DESCRIPTOR_SET_UNIFORM_BUFFERS_DYNAMIC +
+		MAX_DESCRIPTOR_SET_STORAGE_BUFFERS_DYNAMIC,
+	MAX_PUSH_CONSTANT_SIZE = 128,
+
+	MIN_STORAGE_BUFFER_OFFSET_ALIGNMENT = 256,
+
+	REQUIRED_MEMORY_ALIGNMENT = 16,
+
+	SIMD_WIDTH = 4,
+};
+
+struct image_descriptor {
+	ALIGN(16, void *ptr);
+	int width;
+	int height;
+	int depth;
+	int row_pitch_bytes;
+	int slice_pitch_bytes;
+	int sample_pitch_bytes;
+	int sample_count;
+	int size_in_bytes;
+
+	void *stencil_ptr;
+	int stencil_row_pitch_bytes;
+	int stencil_slice_pitch_bytes;
+	int stencil_sample_pitch_bytes;
+
+	// TODO: unused?
+	void *memoryOwner;
+};
+
+struct buffer_descriptor {
+	ALIGN(16, void *ptr);
+	int size_in_bytes;
+	int robustness_size;
+};
+
+struct program_data {
+	uint8_t *descriptor_sets[MAX_BOUND_DESCRIPTOR_SETS];
+	uint32_t descriptor_dynamic_offsets[MAX_DESCRIPTOR_SET_COMBINED_BUFFERS_DYNAMIC];
+	uint4 num_workgroups;
+	uint4 workgroup_size;
+	uint32_t invocations_per_subgroup;
+	uint32_t subgroups_per_workgroup;
+	uint32_t invocations_per_workgroup;
+	unsigned char push_constants[MAX_PUSH_CONSTANT_SIZE];
+	// Unused.
+	void *constants;
+};
+
+typedef int32_t yield_result;
+
+typedef void * coroutine;
+
+typedef coroutine (*routine_begin)(struct program_data *data,
+	int32_t workgroupX,
+	int32_t workgroupY,
+	int32_t workgroupZ,
+	void *workgroupMemory,
+	int32_t firstSubgroup,
+	int32_t subgroupCount);
+
+typedef bool (*routine_await)(coroutine r, yield_result *res);
+
+typedef void (*routine_destroy)(coroutine r);
+

gio/cpu/driver_nosupport.go

@@ -0,0 +1,61 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package cpu
+
+import "unsafe"
+
+type (
+	BufferDescriptor  struct{}
+	ImageDescriptor   struct{}
+	SamplerDescriptor struct{}
+
+	DispatchContext struct{}
+	ThreadContext   struct{}
+	ProgramInfo     struct{}
+)
+
+const Supported = false
+
+func NewBuffer(size int) BufferDescriptor {
+	panic("unsupported")
+}
+
+func (d *BufferDescriptor) Data() []byte {
+	panic("unsupported")
+}
+
+func (d *BufferDescriptor) Free() {
+}
+
+func NewImageRGBA(width, height int) ImageDescriptor {
+	panic("unsupported")
+}
+
+func (d *ImageDescriptor) Data() []byte {
+	panic("unsupported")
+}
+
+func (d *ImageDescriptor) Free() {
+}
+
+func NewDispatchContext() *DispatchContext {
+	panic("unsupported")
+}
+
+func (c *DispatchContext) Free() {
+}
+
+func (c *DispatchContext) Prepare(numThreads int, prog *ProgramInfo, descSet unsafe.Pointer, x, y, z int) {
+	panic("unsupported")
+}
+
+func (c *DispatchContext) Dispatch(threadIdx int, ctx *ThreadContext) {
+	panic("unsupported")
+}
+
+func NewThreadContext() *ThreadContext {
+	panic("unsupported")
+}
+
+func (c *ThreadContext) Free() {
+}

gio/cpu/embed.go

@@ -0,0 +1,11 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package cpu
+
+import _ "embed"
+
+//go:embed abi.h
+var ABIH []byte
+
+//go:embed runtime.h
+var RuntimeH []byte

gio/cpu/go.mod

@@ -0,0 +1,3 @@
+module github.com/empathicqubit/giouibind/gio/cpu
+
+go 1.17

gio/cpu/runtime.h

@@ -0,0 +1,45 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#define ATTR_HIDDEN __attribute__ ((visibility ("hidden")))
+
+// program_info contains constant parameters for a program.
+struct program_info {
+	// MinMemorySize is the minimum size of memory passed to dispatch.
+	size_t min_memory_size;
+	// has_cbarriers is 1 when the program contains control barriers.
+	bool has_cbarriers;
+	// desc_set_size is the size of the first descriptor set for the program.
+	size_t desc_set_size;
+	int workgroup_size_x;
+	int workgroup_size_y;
+	int workgroup_size_z;
+	// Program entrypoints.
+	routine_begin begin;
+	routine_await await;
+	routine_destroy destroy;
+};
+
+// dispatch_context contains the information a program dispatch.
+struct dispatch_context;
+
+// thread_context contains the working memory of a batch. It may be
+// reused, but not concurrently.
+struct thread_context;
+
+extern struct buffer_descriptor alloc_buffer(size_t size) ATTR_HIDDEN;
+extern struct image_descriptor alloc_image_rgba(int width, int height) ATTR_HIDDEN;
+
+extern struct dispatch_context *alloc_dispatch_context(void) ATTR_HIDDEN;
+
+extern void free_dispatch_context(struct dispatch_context *c) ATTR_HIDDEN;
+
+extern struct thread_context *alloc_thread_context(void) ATTR_HIDDEN;
+
+extern void free_thread_context(struct thread_context *c) ATTR_HIDDEN;
+
+// prepare_dispatch initializes ctx to run a dispatch of a program distributed
+// among nthreads threads.
+extern void prepare_dispatch(struct dispatch_context *ctx, int nthreads, struct program_info *info, uint8_t *desc_set, int ngroupx, int ngroupy, int ngroupz) ATTR_HIDDEN;
+
+// dispatch_batch executes a dispatch batch.
+extern void dispatch_thread(struct dispatch_context *ctx, int thread_idx, struct thread_context *thread) ATTR_HIDDEN;

gio/shader/LICENSE

@@ -0,0 +1,63 @@
+This project is provided under the terms of the UNLICENSE or
+the MIT license denoted by the following SPDX identifier:
+
+SPDX-License-Identifier: Unlicense OR MIT
+
+You may use the project under the terms of either license.
+
+Both licenses are reproduced below.
+
+----
+The MIT License (MIT)
+
+Copyright (c) 2019 The Gio authors
+
+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.
+---
+
+
+
+---
+The UNLICENSE
+
+This is free and unencumbered software released into the public domain.
+
+Anyone is free to copy, modify, publish, use, compile, sell, or
+distribute this software, either in source code form or as a compiled
+binary, for any purpose, commercial or non-commercial, and by any
+means.
+
+In jurisdictions that recognize copyright laws, the author or authors
+of this software dedicate any and all copyright interest in the
+software to the public domain. We make this dedication for the benefit
+of the public at large and to the detriment of our heirs and
+successors. We intend this dedication to be an overt act of
+relinquishment in perpetuity of all present and future rights to this
+software under copyright law.
+
+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 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.
+
+For more information, please refer to <https://unlicense.org/>
+---

gio/shader/README.md

@@ -0,0 +1,18 @@
+# GPU programs for the Gio project
+
+This repository contains the source code for the [Gio](https://gioui.org)
+project. It also contains the generators and dereived versions for use with the
+GPU APIs supported by Gio.
+
+# Generating CPU fallbacks
+
+The `piet/gencpu.sh` script updates the piet-gpu binaries:
+
+```
+$ cd piet
+$ ./gencpu.sh
+```
+
+## Issues and contributions
+
+See the [Gio contribution guide](https://gioui.org/doc/contribute).

gio/shader/cmd/convertshaders/glslvalidate.go

@@ -0,0 +1,64 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package main
+
+import (
+	"bytes"
+	"fmt"
+	"io/ioutil"
+	"os/exec"
+	"path/filepath"
+	"strings"
+)
+
+// GLSLValidator is OpenGL reference compiler.
+type GLSLValidator struct {
+	Bin     string
+	WorkDir WorkDir
+}
+
+func NewGLSLValidator() *GLSLValidator { return &GLSLValidator{Bin: "glslangValidator"} }
+
+// Convert converts a glsl shader to spirv.
+func (glsl *GLSLValidator) Convert(path, variant string, lang string, input []byte) ([]byte, error) {
+	base := glsl.WorkDir.Path(filepath.Base(path), variant)
+	pathout := base + ".out"
+
+	cmd := exec.Command(glsl.Bin,
+		"--stdin",
+		"-DLANG_"+strings.ToUpper(lang),
+		"-I"+filepath.Dir(path),
+		"-V", // OpenGL ES 3.1.
+		"-w", // Suppress warnings.
+		"-S", filepath.Ext(path)[1:],
+		"-o", pathout,
+	)
+	cmd.Stdin = bytes.NewBuffer(input)
+
+	out, err := cmd.Output()
+	if err != nil {
+		return nil, fmt.Errorf("%s\nfailed to run %v: %w", out, cmd.Args, err)
+	}
+
+	compiled, err := ioutil.ReadFile(pathout)
+	if err != nil {
+		return nil, fmt.Errorf("unable to read output %q: %w", pathout, err)
+	}
+
+	return compiled, nil
+}
+
+func spirvOpt(spirv []byte) ([]byte, error) {
+	cmd := exec.Command("spirv-opt",
+		"-O",
+		"-",
+		"-o", "-",
+	)
+	cmd.Stdin = bytes.NewBuffer(spirv)
+
+	out, err := cmd.Output()
+	if err != nil {
+		return nil, fmt.Errorf("%s\nfailed to run %v: %w", out, cmd.Args, err)
+	}
+	return out, nil
+}

gio/shader/cmd/convertshaders/hlsl.go

@@ -0,0 +1,146 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package main
+
+import (
+	"bytes"
+	"fmt"
+	"io"
+	"io/ioutil"
+	"os/exec"
+	"path/filepath"
+	"runtime"
+	"strings"
+)
+
+// FXC is hlsl compiler that targets ShaderModel 5.x and lower.
+type FXC struct {
+	Bin     string
+	WorkDir WorkDir
+}
+
+func NewFXC() *FXC { return &FXC{Bin: "fxc.exe"} }
+
+// Compile compiles the input shader.
+func (fxc *FXC) Compile(path, variant string, input []byte, entryPoint string, profileVersion string) ([]byte, error) {
+	base := fxc.WorkDir.Path(filepath.Base(path), variant, profileVersion)
+	pathin := base + ".in"
+	pathout := base + ".out"
+	result := pathout
+
+	if err := fxc.WorkDir.WriteFile(pathin, input); err != nil {
+		return nil, fmt.Errorf("unable to write shader to disk: %w", err)
+	}
+
+	cmd := exec.Command(fxc.Bin)
+	if runtime.GOOS != "windows" {
+		cmd = exec.Command("wine", fxc.Bin)
+		if err := winepath(&pathin, &pathout); err != nil {
+			return nil, err
+		}
+	}
+
+	var profile string
+	switch filepath.Ext(path) {
+	case ".frag":
+		profile = "ps_" + profileVersion
+	case ".vert":
+		profile = "vs_" + profileVersion
+	case ".comp":
+		profile = "cs_" + profileVersion
+	default:
+		return nil, fmt.Errorf("unrecognized shader type %s", path)
+	}
+
+	cmd.Args = append(cmd.Args,
+		"/Fo", pathout,
+		"/T", profile,
+		"/E", entryPoint,
+		pathin,
+	)
+
+	output, err := cmd.CombinedOutput()
+	if err != nil {
+		info := ""
+		if runtime.GOOS != "windows" {
+			info = "If the fxc tool cannot be found, set WINEPATH to the Windows path for the Windows SDK.\n"
+		}
+		return nil, fmt.Errorf("%s\n%sfailed to run %v: %w", output, info, cmd.Args, err)
+	}
+
+	compiled, err := ioutil.ReadFile(result)
+	if err != nil {
+		return nil, fmt.Errorf("unable to read output %q: %w", pathout, err)
+	}
+
+	return compiled, nil
+}
+
+// DXC is hlsl compiler that targets ShaderModel 6.0 and newer.
+type DXC struct {
+	Bin     string
+	WorkDir WorkDir
+}
+
+func NewDXC() *DXC { return &DXC{Bin: "dxc"} }
+
+// Compile compiles the input shader.
+func (dxc *DXC) Compile(path, variant string, input []byte, entryPoint string, profile string) (string, error) {
+	base := dxc.WorkDir.Path(filepath.Base(path), variant, profile)
+	pathin := base + ".in"
+	pathout := base + ".out"
+	result := pathout
+
+	if err := dxc.WorkDir.WriteFile(pathin, input); err != nil {
+		return "", fmt.Errorf("unable to write shader to disk: %w", err)
+	}
+
+	cmd := exec.Command(dxc.Bin)
+
+	cmd.Args = append(cmd.Args,
+		"-Fo", pathout,
+		"-T", profile,
+		"-E", entryPoint,
+		"-Qstrip_reflect",
+		pathin,
+	)
+
+	output, err := cmd.CombinedOutput()
+	if err != nil {
+		return "", fmt.Errorf("%s\nfailed to run %v: %w", output, cmd.Args, err)
+	}
+
+	compiled, err := ioutil.ReadFile(result)
+	if err != nil {
+		return "", fmt.Errorf("unable to read output %q: %w", pathout, err)
+	}
+
+	return string(compiled), nil
+}
+
+// winepath uses the winepath tool to convert a paths to Windows format.
+// The returned path can be used as arguments for Windows command line tools.
+func winepath(paths ...*string) error {
+	winepath := exec.Command("winepath", "--windows")
+	for _, path := range paths {
+		winepath.Args = append(winepath.Args, *path)
+	}
+	// Use a pipe instead of Output, because winepath may have left wineserver
+	// running for several seconds as a grandchild.
+	out, err := winepath.StdoutPipe()
+	if err != nil {
+		return fmt.Errorf("unable to start winepath: %w", err)
+	}
+	if err := winepath.Start(); err != nil {
+		return fmt.Errorf("unable to start winepath: %w", err)
+	}
+	var buf bytes.Buffer
+	if _, err := io.Copy(&buf, out); err != nil {
+		return fmt.Errorf("unable to run winepath: %w", err)
+	}
+	winPaths := strings.Split(strings.TrimSpace(buf.String()), "\n")
+	for i, path := range paths {
+		*path = winPaths[i]
+	}
+	return nil
+}

gio/shader/cmd/convertshaders/main.go

@@ -0,0 +1,547 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package main
+
+import (
+	"bytes"
+	"errors"
+	"flag"
+	"fmt"
+	"io"
+	"io/ioutil"
+	"os"
+	"os/exec"
+	"path/filepath"
+	"sort"
+	"strconv"
+	"strings"
+	"sync"
+	"text/template"
+)
+
+type shaderSources struct {
+	Name      string
+	SPIRV     []byte
+	GLSL100ES []byte
+	GLSL150   []byte
+	DXBC      []byte
+	MetalLibs MetalLibs
+	Reflect   Metadata
+}
+
+func main() {
+	packageName := flag.String("package", "", "specify Go package name")
+	workdir := flag.String("work", "", "temporary working directory (default TEMP)")
+	shadersDir := flag.String("dir", "shaders", "shaders directory")
+
+	flag.Parse()
+
+	var work WorkDir
+	cleanup := func() {}
+	if *workdir == "" {
+		tempdir, err := ioutil.TempDir("", "shader-convert")
+		if err != nil {
+			fmt.Fprintf(os.Stderr, "failed to create tempdir: %v\n", err)
+			os.Exit(1)
+		}
+		cleanup = func() { os.RemoveAll(tempdir) }
+		defer cleanup()
+
+		work = WorkDir(tempdir)
+	} else {
+		if abs, err := filepath.Abs(*workdir); err == nil {
+			*workdir = abs
+		}
+		work = WorkDir(*workdir)
+	}
+
+	var out bytes.Buffer
+	conv := NewConverter(work, *packageName, *shadersDir)
+	if err := conv.Run(&out); err != nil {
+		fmt.Fprintf(os.Stderr, "%v\n", err)
+		cleanup()
+		os.Exit(1)
+	}
+
+	if err := ioutil.WriteFile("shaders.go", out.Bytes(), 0644); err != nil {
+		fmt.Fprintf(os.Stderr, "failed to create shaders: %v\n", err)
+		cleanup()
+		os.Exit(1)
+	}
+
+	cmd := exec.Command("gofmt", "-s", "-w", "shaders.go")
+	cmd.Stdout, cmd.Stderr = os.Stdout, os.Stderr
+	if err := cmd.Run(); err != nil {
+		fmt.Fprintf(os.Stderr, "formatting shaders.go failed: %v\n", err)
+		cleanup()
+		os.Exit(1)
+	}
+}
+
+type Converter struct {
+	workDir    WorkDir
+	shadersDir string
+
+	packageName string
+
+	glslvalidator *GLSLValidator
+	spirv         *SPIRVCross
+	fxc           *FXC
+	msl           *MSL
+}
+
+func NewConverter(workDir WorkDir, packageName, shadersDir string) *Converter {
+	if abs, err := filepath.Abs(shadersDir); err == nil {
+		shadersDir = abs
+	}
+
+	conv := &Converter{}
+	conv.workDir = workDir
+	conv.shadersDir = shadersDir
+
+	conv.packageName = packageName
+
+	conv.glslvalidator = NewGLSLValidator()
+	conv.spirv = NewSPIRVCross()
+	conv.fxc = NewFXC()
+	conv.msl = &MSL{
+		WorkDir: workDir.Dir("msl"),
+	}
+
+	verifyBinaryPath(&conv.glslvalidator.Bin)
+	verifyBinaryPath(&conv.spirv.Bin)
+	// We cannot check fxc nor msl since they may depend on wine.
+
+	conv.glslvalidator.WorkDir = workDir.Dir("glslvalidator")
+	conv.fxc.WorkDir = workDir.Dir("fxc")
+	conv.spirv.WorkDir = workDir.Dir("spirv")
+
+	return conv
+}
+
+func verifyBinaryPath(bin *string) {
+	new, err := exec.LookPath(*bin)
+	if err != nil {
+		fmt.Fprintf(os.Stderr, "unable to find %q: %v\n", *bin, err)
+	} else {
+		*bin = new
+	}
+}
+
+func (conv *Converter) Run(out io.Writer) error {
+	shaders, err := filepath.Glob(filepath.Join(conv.shadersDir, "*"))
+	if err != nil {
+		return fmt.Errorf("failed to list shaders in %q: %w", conv.shadersDir, err)
+	}
+
+	sort.Strings(shaders)
+
+	var workers Workers
+
+	type ShaderResult struct {
+		Path    string
+		Shaders []shaderSources
+		Error   error
+	}
+	shaderResults := make([]ShaderResult, len(shaders))
+
+	for i, shaderPath := range shaders {
+		i, shaderPath := i, shaderPath
+
+		switch filepath.Ext(shaderPath) {
+		case ".vert", ".frag":
+			workers.Go(func() {
+				shaders, err := conv.Shader(shaderPath)
+				shaderResults[i] = ShaderResult{
+					Path:    shaderPath,
+					Shaders: shaders,
+					Error:   err,
+				}
+			})
+		case ".comp":
+			workers.Go(func() {
+				shaders, err := conv.ComputeShader(shaderPath)
+				shaderResults[i] = ShaderResult{
+					Path:    shaderPath,
+					Shaders: shaders,
+					Error:   err,
+				}
+			})
+		default:
+			continue
+		}
+	}
+
+	workers.Wait()
+
+	var allErrors string
+	for _, r := range shaderResults {
+		if r.Error != nil {
+			if len(allErrors) > 0 {
+				allErrors += "\n\n"
+			}
+			allErrors += "--- " + r.Path + " --- \n\n" + r.Error.Error() + "\n"
+		}
+	}
+	if len(allErrors) > 0 {
+		return errors.New(allErrors)
+	}
+
+	fmt.Fprintf(out, "// Code generated by build.go. DO NOT EDIT.\n\n")
+	fmt.Fprintf(out, "package %s\n\n", conv.packageName)
+	fmt.Fprintf(out, "import (\n")
+	fmt.Fprintf(out, "\t%q\n", "runtime")
+	fmt.Fprintf(out, "\t_ %q\n", "embed")
+	fmt.Fprintln(out)
+	fmt.Fprintf(out, "\t%q\n", "gioui.org/shader")
+	fmt.Fprintf(out, ")\n\n")
+
+	fmt.Fprintf(out, "var (\n")
+
+	var genErr error
+	for _, r := range shaderResults {
+		if len(r.Shaders) == 0 {
+			continue
+		}
+
+		name := filepath.Base(r.Path)
+		name = strings.ReplaceAll(name, ".", "_")
+		fmt.Fprintf(out, "\tShader_%s = ", name)
+
+		multiVariant := len(r.Shaders) > 1
+		if multiVariant {
+			fmt.Fprintf(out, "[...]shader.Sources{\n")
+		}
+
+		writeGenerated := func(src []byte, prefix, path string, idx int) {
+			if len(src) == 0 || genErr != nil {
+				return
+			}
+			base := fmt.Sprintf("z%s.%d.%s", filepath.Base(path), idx, strings.ToLower(prefix))
+			p := filepath.Join(filepath.Dir(path), base)
+			genErr = os.WriteFile(p, src, 0o644)
+		}
+		for i, src := range r.Shaders {
+			fmt.Fprintf(out, "shader.Sources{\n")
+			fmt.Fprintf(out, "Name: %#v,\n", src.Name)
+			if inp := src.Reflect.Inputs; len(inp) > 0 {
+				fmt.Fprintf(out, "Inputs: %#v,\n", inp)
+			}
+			if u := src.Reflect.Uniforms; u.Size > 0 {
+				fmt.Fprintf(out, "Uniforms: shader.UniformsReflection{\n")
+				fmt.Fprintf(out, "Locations: %#v,\n", u.Locations)
+				fmt.Fprintf(out, "Size: %d,\n", u.Size)
+				fmt.Fprintf(out, "},\n")
+			}
+			if tex := src.Reflect.Textures; len(tex) > 0 {
+				fmt.Fprintf(out, "Textures: %#v,\n", tex)
+			}
+			if imgs := src.Reflect.Images; len(imgs) > 0 {
+				fmt.Fprintf(out, "Images: %#v,\n", imgs)
+			}
+			if bufs := src.Reflect.StorageBuffers; len(bufs) > 0 {
+				fmt.Fprintf(out, "StorageBuffers: %#v,\n", bufs)
+			}
+			if wg := src.Reflect.WorkgroupSize; wg != [3]int{} {
+				fmt.Fprintf(out, "WorkgroupSize: %#v,\n", wg)
+			}
+			writeGenerated(src.SPIRV, "SPIRV", r.Path, i)
+			writeGenerated(src.GLSL100ES, "GLSL100ES", r.Path, i)
+			writeGenerated(src.GLSL150, "GLSL150", r.Path, i)
+			writeGenerated(src.DXBC, "DXBC", r.Path, i)
+			writeGenerated(src.MetalLibs.MacOS, "MetalLibMacOS", r.Path, i)
+			writeGenerated(src.MetalLibs.IOS, "MetalLibIOS", r.Path, i)
+			writeGenerated(src.MetalLibs.IOSSimulator, "MetalLibIOSSimulator", r.Path, i)
+			fmt.Fprintf(out, "}")
+			if multiVariant {
+				fmt.Fprintf(out, ",")
+			}
+			fmt.Fprintf(out, "\n")
+		}
+		if multiVariant {
+			fmt.Fprintf(out, "}\n")
+		}
+		writeEmbedded := func(src []byte, prefix, path string, idx int) {
+			base := fmt.Sprintf("z%s.%d.%s", filepath.Base(path), idx, strings.ToLower(prefix))
+			if _, err := os.Stat(base); err != nil {
+				return
+			}
+			field := strings.ReplaceAll(base, ".", "_")
+			fmt.Fprintf(out, "//go:embed %s\n", base)
+			fmt.Fprintf(out, "%s string\n", field)
+		}
+		for i, src := range r.Shaders {
+			writeEmbedded(src.SPIRV, "SPIRV", r.Path, i)
+			writeEmbedded(src.GLSL100ES, "GLSL100ES", r.Path, i)
+			writeEmbedded(src.GLSL150, "GLSL150", r.Path, i)
+			writeEmbedded(src.DXBC, "DXBC", r.Path, i)
+			writeEmbedded(src.MetalLibs.MacOS, "MetalLibMacOS", r.Path, i)
+			writeEmbedded(src.MetalLibs.IOS, "MetalLibIOS", r.Path, i)
+			writeEmbedded(src.MetalLibs.IOSSimulator, "MetalLibIOSSimulator", r.Path, i)
+		}
+	}
+	fmt.Fprintf(out, ")\n")
+	writeInit := func(src []byte, prefix, field, path string, idx int, variants bool) {
+		name := filepath.Base(path)
+		name = strings.ReplaceAll(name, ".", "_")
+		base := fmt.Sprintf("z%s.%d.%s", filepath.Base(path), idx, strings.ToLower(prefix))
+		if _, err := os.Stat(base); err != nil {
+			return
+		}
+		variable := strings.ReplaceAll(base, ".", "_")
+		index := ""
+		if variants {
+			index = fmt.Sprintf("[%d]", idx)
+		}
+		fmt.Fprintf(out, "\t\tShader_%s%s.%s = %s\n", name, index, field, variable)
+	}
+	fmt.Fprintf(out, "func init() {\n")
+	fmt.Fprintf(out, "\tconst (\n")
+	fmt.Fprintf(out, "\t\topengles = %s\n", geeseExpr("linux", "freebsd", "openbsd", "windows", "js", "android", "darwin", "ios"))
+	fmt.Fprintf(out, "\t\topengl = %s\n", geeseExpr("darwin"))
+	fmt.Fprintf(out, "\t\td3d11 = %s\n", geeseExpr("windows"))
+	fmt.Fprintf(out, "\t\tvulkan = %s\n", geeseExpr("linux", "android"))
+	fmt.Fprintf(out, "\t)\n")
+	for _, r := range shaderResults {
+		variants := len(r.Shaders) > 1
+		for i, src := range r.Shaders {
+			fmt.Fprintf(out, "\tif vulkan {\n")
+			writeInit(src.SPIRV, "SPIRV", "SPIRV", r.Path, i, variants)
+			fmt.Fprintf(out, "\t}\n")
+			fmt.Fprintf(out, "\tif opengles {\n")
+			writeInit(src.GLSL100ES, "GLSL100ES", "GLSL100ES", r.Path, i, variants)
+			fmt.Fprintf(out, "\t}\n")
+			fmt.Fprintf(out, "\tif opengl {\n")
+			writeInit(src.GLSL150, "GLSL150", "GLSL150", r.Path, i, variants)
+			fmt.Fprintf(out, "\t}\n")
+			fmt.Fprintf(out, "\tif d3d11 {\n")
+			writeInit(src.DXBC, "DXBC", "DXBC", r.Path, i, variants)
+			fmt.Fprintf(out, "\t}\n")
+			fmt.Fprintf(out, "\tif runtime.GOOS == \"darwin\" {\n")
+			writeInit(src.MetalLibs.MacOS, "MetalLibMacOS", "MetalLib", r.Path, i, variants)
+			fmt.Fprintf(out, "\t}\n")
+			fmt.Fprintf(out, "\tif runtime.GOOS == \"ios\" {\n")
+			fmt.Fprintf(out, "if runtime.GOARCH == \"amd64\" {\n")
+			writeInit(src.MetalLibs.IOSSimulator, "MetalLibIOSSimulator", "MetalLib", r.Path, i, variants)
+			fmt.Fprintf(out, "\t\t} else {\n")
+			writeInit(src.MetalLibs.IOS, "MetalLibIOS", "MetalLib", r.Path, i, variants)
+			fmt.Fprintf(out, "\t\t}\n")
+			fmt.Fprintf(out, "\t}\n")
+		}
+	}
+	fmt.Fprintf(out, "}\n")
+
+	return genErr
+}
+
+func geeseExpr(geese ...string) string {
+	var checks []string
+	for _, goos := range geese {
+		checks = append(checks, fmt.Sprintf("runtime.GOOS == %q", goos))
+	}
+	return strings.Join(checks, " || ")
+}
+
+func (conv *Converter) Shader(shaderPath string) ([]shaderSources, error) {
+	type Variant struct {
+		FetchColorExpr string
+		Header         string
+	}
+	variantArgs := [...]Variant{
+		{
+			FetchColorExpr: `_color.color`,
+			Header:         `layout(push_constant) uniform Color { layout(offset=112) vec4 color; } _color;`,
+		},
+		{
+			FetchColorExpr: `mix(_gradient.color1, _gradient.color2, clamp(vUV.x, 0.0, 1.0))`,
+			Header:         `layout(push_constant) uniform Gradient { layout(offset=96) vec4 color1; vec4 color2; } _gradient;`,
+		},
+		{
+			FetchColorExpr: `texture(tex, vUV)`,
+			Header:         `layout(binding=0) uniform sampler2D tex;`,
+		},
+	}
+
+	shaderTemplate, err := template.ParseFiles(shaderPath)
+	if err != nil {
+		return nil, fmt.Errorf("failed to parse template %q: %w", shaderPath, err)
+	}
+
+	var variants []shaderSources
+	for i, variantArg := range variantArgs {
+		variantName := strconv.Itoa(i)
+		var buf bytes.Buffer
+		err := shaderTemplate.Execute(&buf, variantArg)
+		if err != nil {
+			return nil, fmt.Errorf("failed to execute template %q with %#v: %w", shaderPath, variantArg, err)
+		}
+
+		var sources shaderSources
+		sources.Name = filepath.Base(shaderPath)
+
+		src := buf.Bytes()
+		sources.SPIRV, err = conv.glslvalidator.Convert(shaderPath, variantName, "vulkan", src)
+		if err != nil {
+			return nil, fmt.Errorf("failed to generate SPIR-V for %q: %w", shaderPath, err)
+		}
+
+		sources.SPIRV, err = spirvOpt(sources.SPIRV)
+		if err != nil {
+			return nil, fmt.Errorf("failed to optimize SPIR-V for %q: %w", shaderPath, err)
+		}
+
+		var reflect Metadata
+		sources.GLSL100ES, reflect, err = conv.ShaderVariant(shaderPath, variantName, src, "es", "100")
+		if err != nil {
+			return nil, fmt.Errorf("failed to convert GLSL100ES:\n%w", err)
+		}
+
+		metal, _, err := conv.ShaderVariant(shaderPath, variantName, src, "msl", "10000")
+		if err != nil {
+			return nil, fmt.Errorf("failed to convert to Metal:\n%w", err)
+		}
+
+		metalIOS, _, err := conv.ShaderVariant(shaderPath, variantName, src, "mslios", "10000")
+		if err != nil {
+			return nil, fmt.Errorf("failed to convert to Metal:\n%w", err)
+		}
+
+		sources.MetalLibs, err = conv.msl.Compile(shaderPath, variantName, metal, metalIOS)
+		if err != nil {
+			if !errors.Is(err, exec.ErrNotFound) {
+				return nil, fmt.Errorf("failed to build .metallib library:\n%w", err)
+			}
+		}
+
+		hlsl, _, err := conv.ShaderVariant(shaderPath, variantName, src, "hlsl", "40")
+		if err != nil {
+			return nil, fmt.Errorf("failed to convert HLSL:\n%w", err)
+		}
+
+		sources.DXBC, err = conv.fxc.Compile(shaderPath, variantName, []byte(hlsl), "main", "4_0_level_9_1")
+		if err != nil {
+			// Attempt shader model 4.0. Only the gpu/headless
+			// test shaders use features not supported by level
+			// 9.1.
+			sources.DXBC, err = conv.fxc.Compile(shaderPath, variantName, []byte(hlsl), "main", "4_0")
+			if err != nil {
+				if !errors.Is(err, exec.ErrNotFound) {
+					return nil, fmt.Errorf("failed to compile HLSL: %w", err)
+				}
+			}
+		}
+
+		sources.GLSL150, _, err = conv.ShaderVariant(shaderPath, variantName, src, "glsl", "150")
+		if err != nil {
+			return nil, fmt.Errorf("failed to convert GLSL150:\n%w", err)
+		}
+
+		sources.Reflect = reflect
+
+		variants = append(variants, sources)
+	}
+
+	// If the shader don't use the variant arguments, output only a single version.
+	if bytes.Equal(variants[0].GLSL100ES, variants[1].GLSL100ES) {
+		variants = variants[:1]
+	}
+
+	return variants, nil
+}
+
+func (conv *Converter) ShaderVariant(shaderPath, variant string, src []byte, lang, profile string) ([]byte, Metadata, error) {
+	spirv, err := conv.glslvalidator.Convert(shaderPath, variant, lang, src)
+	if err != nil {
+		return nil, Metadata{}, fmt.Errorf("failed to generate SPIR-V for %q: %w", shaderPath, err)
+	}
+
+	dst, err := conv.spirv.Convert(shaderPath, variant, spirv, lang, profile)
+	if err != nil {
+		return nil, Metadata{}, fmt.Errorf("failed to convert shader %q: %w", shaderPath, err)
+	}
+
+	meta, err := conv.spirv.Metadata(shaderPath, variant, spirv)
+	if err != nil {
+		return nil, Metadata{}, fmt.Errorf("failed to extract metadata for shader %q: %w", shaderPath, err)
+	}
+
+	return dst, meta, nil
+}
+
+func (conv *Converter) ComputeShader(shaderPath string) ([]shaderSources, error) {
+	sh, err := ioutil.ReadFile(shaderPath)
+	if err != nil {
+		return nil, fmt.Errorf("failed to load shader %q: %w", shaderPath, err)
+	}
+
+	sources := shaderSources{
+		Name: filepath.Base(shaderPath),
+	}
+	spirv, err := conv.glslvalidator.Convert(shaderPath, "", "glsl", sh)
+	if err != nil {
+		return nil, fmt.Errorf("failed to convert compute shader %q: %w", shaderPath, err)
+	}
+
+	sources.SPIRV, err = spirvOpt(spirv)
+	if err != nil {
+		return nil, fmt.Errorf("failed to optimize SPIR-V for %q: %w", shaderPath, err)
+	}
+
+	meta, err := conv.spirv.Metadata(shaderPath, "", spirv)
+	if err != nil {
+		return nil, fmt.Errorf("failed to extract metadata for shader %q: %w", shaderPath, err)
+	}
+	sources.Reflect = meta
+
+	metal, err := conv.spirv.Convert(shaderPath, "", spirv, "msl", "10000")
+	if err != nil {
+		return nil, fmt.Errorf("failed to convert GLSL130:\n%w", err)
+	}
+
+	metalIOS, err := conv.spirv.Convert(shaderPath, "", spirv, "mslios", "10000")
+	if err != nil {
+		return nil, fmt.Errorf("failed to convert GLSL130:\n%w", err)
+	}
+
+	sources.MetalLibs, err = conv.msl.Compile(shaderPath, "", metal, metalIOS)
+	if err != nil {
+		if !errors.Is(err, exec.ErrNotFound) {
+			return nil, fmt.Errorf("failed to build .metallib library:\n%w", err)
+		}
+	}
+
+	hlslSource, err := conv.spirv.Convert(shaderPath, "", spirv, "hlsl", "50")
+	if err != nil {
+		return nil, fmt.Errorf("failed to convert hlsl compute shader %q: %w", shaderPath, err)
+	}
+
+	sources.DXBC, err = conv.fxc.Compile(shaderPath, "0", []byte(hlslSource), "main", "5_0")
+	if err != nil {
+		if !errors.Is(err, exec.ErrNotFound) {
+			return nil, fmt.Errorf("failed to compile hlsl compute shader %q: %w", shaderPath, err)
+		}
+	}
+
+	return []shaderSources{sources}, nil
+}
+
+// Workers implements wait group with synchronous logging.
+type Workers struct {
+	running sync.WaitGroup
+}
+
+func (lg *Workers) Go(fn func()) {
+	lg.running.Add(1)
+	go func() {
+		defer lg.running.Done()
+		fn()
+	}()
+}
+
+func (lg *Workers) Wait() {
+	lg.running.Wait()
+}
+
+func unixLineEnding(s []byte) []byte {
+	return bytes.ReplaceAll(s, []byte("\r\n"), []byte("\n"))
+}

gio/shader/cmd/convertshaders/msl.go

@@ -0,0 +1,108 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package main
+
+import (
+	"fmt"
+	"io/ioutil"
+	"os"
+	"os/exec"
+	"path/filepath"
+	"runtime"
+)
+
+// MSL is hlsl compiler that targets the Metal shading language
+type MSL struct {
+	WorkDir WorkDir
+}
+
+// MetalLibs contains compiled .metallib programs for all supported platforms.
+type MetalLibs struct {
+	MacOS        []byte
+	IOS          []byte
+	IOSSimulator []byte
+}
+
+// Compile compiles the input .metal program and converts it into .metallib libraries.
+func (msl *MSL) Compile(path, variant string, src, srcIOS []byte) (MetalLibs, error) {
+	base := msl.WorkDir.Path(filepath.Base(path), variant)
+	pathinMacOS := base + ".macos.metal"
+	pathinIOS := base + ".ios.metal"
+
+	var libs MetalLibs
+	if err := msl.WorkDir.WriteFile(pathinMacOS, []byte(src)); err != nil {
+		return libs, fmt.Errorf("unable to write shader to disk: %w", err)
+	}
+	if err := msl.WorkDir.WriteFile(pathinIOS, []byte(srcIOS)); err != nil {
+		return libs, fmt.Errorf("unable to write shader to disk: %w", err)
+	}
+
+	var err error
+	libs.MacOS, err = msl.compileFor("macosx", "-mmacosx-version-min=10.11", pathinMacOS)
+	if err != nil {
+		return libs, err
+	}
+	libs.IOS, err = msl.compileFor("iphoneos", "-mios-version-min=10.0", pathinIOS)
+	if err != nil {
+		return libs, err
+	}
+	libs.IOSSimulator, err = msl.compileFor("iphonesimulator", "-miphonesimulator-version-min=8.0", pathinIOS)
+	if err != nil {
+		return libs, err
+	}
+
+	return libs, nil
+}
+
+// compileFor compiles the input .metal program and converts it into a
+// .metallib library for a particular SDK.
+func (msl *MSL) compileFor(sdk, minVer, path string) ([]byte, error) {
+	var metal *exec.Cmd
+
+	pathout := path + ".metallib"
+	result := pathout
+
+	if runtime.GOOS == "darwin" {
+		metal = exec.Command("xcrun", "--sdk", sdk, "metal")
+	} else {
+		sdkDir := os.Getenv("METAL_SDK_ROOT")
+		if sdkDir == "" {
+			return nil, exec.ErrNotFound
+		}
+		switch sdk {
+		case "macosx":
+			sdkDir = filepath.Join(sdkDir, "macos")
+		case "iphoneos", "iphonesimulator":
+			sdkDir = filepath.Join(sdkDir, "ios")
+		default:
+			panic("unknown sdk")
+		}
+		bin := filepath.Join(sdkDir, "bin", "metal.exe")
+		if runtime.GOOS == "windows" {
+			metal = exec.Command(bin)
+		} else {
+			if err := winepath(&path, &pathout); err != nil {
+				return nil, err
+			}
+			metal = exec.Command("wine", bin)
+		}
+	}
+
+	metal.Args = append(metal.Args,
+		minVer,
+		"-o", pathout,
+		path,
+	)
+
+	output, err := metal.CombinedOutput()
+	if err != nil {
+		return nil, fmt.Errorf("%s\nfailed to run %v: %w", output, metal.Args, err)
+	}
+
+	compiled, err := ioutil.ReadFile(result)
+	if err != nil {
+		return nil, fmt.Errorf("unable to read output %q: %w", pathout, err)
+	}
+
+	return compiled, nil
+}

gio/shader/cmd/convertshaders/spirvcross.go

@@ -0,0 +1,259 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package main
+
+import (
+	"encoding/json"
+	"fmt"
+	"math"
+	"os/exec"
+	"path/filepath"
+	"sort"
+
+	"gioui.org/shader"
+)
+
+// Metadata contains reflection data about a shader.
+type Metadata struct {
+	Uniforms       shader.UniformsReflection
+	Inputs         []shader.InputLocation
+	Textures       []shader.TextureBinding
+	Images         []shader.ImageBinding
+	StorageBuffers []shader.BufferBinding
+	WorkgroupSize  [3]int
+}
+
+// SPIRVCross cross-compiles spirv shaders to es, hlsl and others.
+type SPIRVCross struct {
+	Bin     string
+	WorkDir WorkDir
+}
+
+func NewSPIRVCross() *SPIRVCross { return &SPIRVCross{Bin: "spirv-cross"} }
+
+// Convert converts compute shader from spirv format to a target format.
+func (spirv *SPIRVCross) Convert(path, variant string, shader []byte, target, version string) ([]byte, error) {
+	base := spirv.WorkDir.Path(filepath.Base(path), variant)
+
+	if err := spirv.WorkDir.WriteFile(base, shader); err != nil {
+		return nil, fmt.Errorf("unable to write shader to disk: %w", err)
+	}
+
+	var cmd *exec.Cmd
+	switch target {
+	case "glsl":
+		cmd = exec.Command(spirv.Bin,
+			"--no-es",
+			"--version", version,
+		)
+	case "es":
+		cmd = exec.Command(spirv.Bin,
+			"--es",
+			"--version", version,
+		)
+	case "hlsl":
+		cmd = exec.Command(spirv.Bin,
+			"--hlsl",
+			"--shader-model", version,
+		)
+	case "msl", "mslios":
+		cmd = exec.Command(spirv.Bin,
+			"--msl",
+			"--msl-decoration-binding",
+			"--msl-version", version,
+		)
+		if target == "mslios" {
+			cmd.Args = append(cmd.Args, "--msl-ios")
+		}
+	default:
+		return nil, fmt.Errorf("unknown target %q", target)
+	}
+	cmd.Args = append(cmd.Args, "--no-420pack-extension", base)
+
+	out, err := cmd.CombinedOutput()
+	if err != nil {
+		return nil, fmt.Errorf("%s\nfailed to run %v: %w", out, cmd.Args, err)
+	}
+	if target != "hlsl" {
+		// Strip Windows \r in line endings.
+		out = unixLineEnding(out)
+	}
+
+	return out, nil
+}
+
+// Metadata extracts metadata for a SPIR-V shader.
+func (spirv *SPIRVCross) Metadata(path, variant string, shader []byte) (Metadata, error) {
+	base := spirv.WorkDir.Path(filepath.Base(path), variant)
+
+	if err := spirv.WorkDir.WriteFile(base, shader); err != nil {
+		return Metadata{}, fmt.Errorf("unable to write shader to disk: %w", err)
+	}
+
+	cmd := exec.Command(spirv.Bin,
+		base,
+		"--reflect",
+	)
+
+	out, err := cmd.Output()
+	if err != nil {
+		return Metadata{}, fmt.Errorf("failed to run %v: %w", cmd.Args, err)
+	}
+
+	meta, err := parseMetadata(out)
+	if err != nil {
+		return Metadata{}, fmt.Errorf("%s\nfailed to parse metadata: %w", out, err)
+	}
+
+	return meta, nil
+}
+
+func parseMetadata(data []byte) (Metadata, error) {
+	var reflect struct {
+		Types map[string]struct {
+			Name    string `json:"name"`
+			Members []struct {
+				Name   string `json:"name"`
+				Type   string `json:"type"`
+				Offset int    `json:"offset"`
+			} `json:"members"`
+		} `json:"types"`
+		Inputs []struct {
+			Name     string `json:"name"`
+			Type     string `json:"type"`
+			Location int    `json:"location"`
+		} `json:"inputs"`
+		Textures []struct {
+			Name    string `json:"name"`
+			Type    string `json:"type"`
+			Set     int    `json:"set"`
+			Binding int    `json:"binding"`
+		} `json:"textures"`
+		UBOs []struct {
+			Name      string `json:"name"`
+			Type      string `json:"type"`
+			BlockSize int    `json:"block_size"`
+			Set       int    `json:"set"`
+			Binding   int    `json:"binding"`
+		} `json:"ubos"`
+		PushConstants []struct {
+			Name         string `json:"name"`
+			Type         string `json:"type"`
+			PushConstant bool   `json:"push_constant"`
+		} `json:"push_constants"`
+		EntryPoints []struct {
+			Name          string `json:"name"`
+			Mode          string `json:"mode"`
+			WorkgroupSize [3]int `json:"workgroup_size"`
+		} `json:"entryPoints"`
+		StorageBuffers []struct {
+			Name    string `json:"name"`
+			Type    string `json:"type"`
+			Binding int    `json:"binding"`
+		} `json:"ssbos"`
+		Images []struct {
+			Name    string `json:"name"`
+			Type    string `json:"type"`
+			Binding int    `json:"binding"`
+		} `json:"images"`
+	}
+	if err := json.Unmarshal(data, &reflect); err != nil {
+		return Metadata{}, fmt.Errorf("failed to parse reflection data: %w", err)
+	}
+
+	var m Metadata
+
+	for _, input := range reflect.Inputs {
+		dataType, dataSize, err := parseDataType(input.Type)
+		if err != nil {
+			return Metadata{}, fmt.Errorf("parseReflection: %v", err)
+		}
+		m.Inputs = append(m.Inputs, shader.InputLocation{
+			Name:          input.Name,
+			Location:      input.Location,
+			Semantic:      "TEXCOORD",
+			SemanticIndex: input.Location,
+			Type:          dataType,
+			Size:          dataSize,
+		})
+	}
+
+	sort.Slice(m.Inputs, func(i, j int) bool {
+		return m.Inputs[i].Location < m.Inputs[j].Location
+	})
+
+	blockSize := 0
+	minOffset := math.MaxInt
+	for _, block := range reflect.PushConstants {
+		t := reflect.Types[block.Type]
+		for _, member := range t.Members {
+			dataType, size, err := parseDataType(member.Type)
+			if err != nil {
+				return Metadata{}, fmt.Errorf("failed to parse reflection data: %v", err)
+			}
+			blockSize += size * 4
+			if member.Offset < minOffset {
+				minOffset = member.Offset
+			}
+			m.Uniforms.Locations = append(m.Uniforms.Locations, shader.UniformLocation{
+				Name:   fmt.Sprintf("%s.%s", block.Name, member.Name),
+				Type:   dataType,
+				Size:   size,
+				Offset: member.Offset,
+			})
+		}
+	}
+	m.Uniforms.Size = blockSize
+
+	for _, texture := range reflect.Textures {
+		m.Textures = append(m.Textures, shader.TextureBinding{
+			Name:    texture.Name,
+			Binding: texture.Binding,
+		})
+	}
+
+	for _, img := range reflect.Images {
+		m.Images = append(m.Images, shader.ImageBinding{
+			Name:    img.Name,
+			Binding: img.Binding,
+		})
+	}
+
+	for _, sb := range reflect.StorageBuffers {
+		m.StorageBuffers = append(m.StorageBuffers, shader.BufferBinding{
+			Name:    sb.Name,
+			Binding: sb.Binding,
+		})
+	}
+
+	for _, e := range reflect.EntryPoints {
+		if e.Name == "main" && e.Mode == "comp" {
+			m.WorkgroupSize = e.WorkgroupSize
+		}
+	}
+
+	return m, nil
+}
+
+func parseDataType(t string) (shader.DataType, int, error) {
+	switch t {
+	case "float":
+		return shader.DataTypeFloat, 1, nil
+	case "vec2":
+		return shader.DataTypeFloat, 2, nil
+	case "vec3":
+		return shader.DataTypeFloat, 3, nil
+	case "vec4":
+		return shader.DataTypeFloat, 4, nil
+	case "int":
+		return shader.DataTypeInt, 1, nil
+	case "int2":
+		return shader.DataTypeInt, 2, nil
+	case "int3":
+		return shader.DataTypeInt, 3, nil
+	case "int4":
+		return shader.DataTypeInt, 4, nil
+	default:
+		return 0, 0, fmt.Errorf("unsupported input data type: %s", t)
+	}
+}

gio/shader/cmd/convertshaders/workdir.go

@@ -0,0 +1,35 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package main
+
+import (
+	"fmt"
+	"io/ioutil"
+	"os"
+	"path/filepath"
+	"strings"
+)
+
+type WorkDir string
+
+func (wd WorkDir) Dir(path string) WorkDir {
+	dirname := filepath.Join(string(wd), path)
+	if err := os.Mkdir(dirname, 0755); err != nil {
+		if !os.IsExist(err) {
+			fmt.Fprintf(os.Stderr, "failed to create %q: %v\n", dirname, err)
+		}
+	}
+	return WorkDir(dirname)
+}
+
+func (wd WorkDir) Path(path ...string) (fullpath string) {
+	return filepath.Join(string(wd), strings.Join(path, "."))
+}
+
+func (wd WorkDir) WriteFile(path string, data []byte) error {
+	err := ioutil.WriteFile(path, data, 0644)
+	if err != nil {
+		return fmt.Errorf("unable to create %v: %w", path, err)
+	}
+	return nil
+}

gio/shader/gio/blit.frag

@@ -0,0 +1,15 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+precision mediump float;
+
+layout(location=0) in highp vec2 vUV;
+
+{{.Header}}
+
+layout(location = 0) out vec4 fragColor;
+
+void main() {
+	fragColor = {{.FetchColorExpr}};
+}

gio/shader/gio/blit.vert

@@ -0,0 +1,27 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#extension GL_GOOGLE_include_directive : enable
+
+precision highp float;
+
+#include "common.h"
+
+layout(push_constant) uniform Block {
+	vec4 transform;
+	vec4 uvTransformR1;
+	vec4 uvTransformR2;
+} _block;
+
+layout(location = 0) in vec2 pos;
+
+layout(location = 1) in vec2 uv;
+
+layout(location = 0) out vec2 vUV;
+
+void main() {
+	vec2 p = pos*_block.transform.xy + _block.transform.zw;
+	gl_Position = vec4(transform3x2(windowTransform, vec3(p, 0)), 1);
+	vUV = transform3x2(m3x2(_block.uvTransformR1.xyz, _block.uvTransformR2.xyz), vec3(uv,1)).xy;
+}

gio/shader/gio/common.h

@@ -0,0 +1,35 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+struct m3x2 {
+	vec3 r0;
+	vec3 r1;
+};
+
+// fboTransform is the transformation that cancels the implied transformation
+// between the clip space and the framebuffer. Only two rows are returned. The
+// last is implied to be [0, 0, 1].
+const m3x2 fboTransform = m3x2(
+#if defined(LANG_HLSL) || defined(LANG_MSL) || defined(LANG_MSLIOS)
+	vec3(1.0, 0.0, 0.0),
+	vec3(0.0, -1.0, 0.0)
+#else
+	vec3(1.0, 0.0, 0.0),
+	vec3(0.0, 1.0, 0.0)
+#endif
+);
+
+// windowTransform is the transformation that cancels the implied transformation
+// between framebuffer space and window system coordinates.
+const m3x2 windowTransform = m3x2(
+#if defined(LANG_VULKAN)
+	vec3(1.0, 0.0, 0.0),
+	vec3(0.0, 1.0, 0.0)
+#else
+	vec3(1.0, 0.0, 0.0),
+	vec3(0.0, -1.0, 0.0)
+#endif
+);
+
+vec3 transform3x2(m3x2 t, vec3 v) {
+	return vec3(dot(t.r0, v), dot(t.r1, v), dot(vec3(0.0, 0.0, 1.0), v));
+}

gio/shader/gio/copy.frag

@@ -0,0 +1,24 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+precision mediump float;
+
+layout(binding = 0) uniform sampler2D tex;
+
+layout(location = 0) in highp vec2 vUV;
+
+layout(location = 0) out vec4 fragColor;
+
+vec3 sRGBtoRGB(vec3 rgb) {
+	bvec3 cutoff = greaterThanEqual(rgb, vec3(0.04045));
+	vec3 below = rgb/vec3(12.92);
+	vec3 above = pow((rgb + vec3(0.055))/vec3(1.055), vec3(2.4));
+	return mix(below, above, cutoff);
+}
+
+void main() {
+	vec4 texel = texture(tex, vUV);
+	texel.rgb = sRGBtoRGB(texel.rgb);
+	fragColor = texel;
+}

gio/shader/gio/copy.vert

@@ -0,0 +1,26 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#extension GL_GOOGLE_include_directive : enable
+
+precision highp float;
+
+#include "common.h"
+
+layout(push_constant) uniform Block {
+	vec2 scale;
+	vec2 pos;
+	vec2 uvScale;
+} _block;
+
+layout(location = 0) in vec2 pos;
+layout(location = 1) in vec2 uv;
+
+layout(location = 0) out vec2 vUV;
+
+void main() {
+	vUV = vec2(uv*_block.uvScale);
+	vec2 p = vec2(pos*_block.scale + _block.pos);
+	gl_Position = vec4(transform3x2(windowTransform, vec3(p, 0)), 1);
+}

gio/shader/gio/cover.frag

@@ -0,0 +1,20 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+precision mediump float;
+
+{{.Header}}
+
+layout(location = 0) in highp vec2 vCoverUV;
+layout(location = 1) in highp vec2 vUV;
+
+layout(binding = 1) uniform sampler2D cover;
+
+layout(location = 0) out vec4 fragColor;
+
+void main() {
+    fragColor = {{.FetchColorExpr}};
+	float c = min(abs(texture(cover, vCoverUV).r), 1.0);
+	fragColor *= c;
+}

gio/shader/gio/cover.vert

@@ -0,0 +1,31 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#extension GL_GOOGLE_include_directive : enable
+
+precision highp float;
+
+#include "common.h"
+
+layout(push_constant) uniform Block {
+	vec4 transform;
+	vec4 uvCoverTransform;
+	vec4 uvTransformR1;
+	vec4 uvTransformR2;
+} _block;
+
+layout(location = 0) in vec2 pos;
+
+layout(location = 0) out vec2 vCoverUV;
+
+layout(location = 1) in vec2 uv;
+layout(location = 1) out vec2 vUV;
+
+void main() {
+	vec2 p = vec2(pos*_block.transform.xy + _block.transform.zw);
+    gl_Position = vec4(transform3x2(windowTransform, vec3(p, 0)), 1);
+	vUV = transform3x2(m3x2(_block.uvTransformR1.xyz, _block.uvTransformR2.xyz), vec3(uv,1)).xy;
+	vec3 uv3 = vec3(uv, 1.0);
+	vCoverUV = (uv3*vec3(_block.uvCoverTransform.xy, 1.0)+vec3(_block.uvCoverTransform.zw, 0.0)).xy;
+}

gio/shader/gio/gen.go

@@ -0,0 +1,5 @@
+// SPDX-License-Identifier: Unlicense OR MIT
+
+package gio
+
+//go:generate go run ../cmd/convertshaders -package gio -dir .

gio/shader/gio/input.vert

@@ -0,0 +1,15 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#extension GL_GOOGLE_include_directive : enable
+
+precision highp float;
+
+#include "common.h"
+
+layout(location=0) in vec4 position;
+
+void main() {
+    gl_Position = vec4(transform3x2(windowTransform, position.xyz), position.w);
+}

gio/shader/gio/intersect.frag

@@ -0,0 +1,15 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+precision mediump float;
+
+layout(location = 0) in highp vec2 vUV;
+
+layout(binding = 0) uniform sampler2D cover;
+
+layout(location = 0) out vec4 fragColor;
+
+void main() {
+  fragColor.r = abs(texture(cover, vUV).r);
+}

gio/shader/gio/intersect.vert

@@ -0,0 +1,26 @@
+#version 310 es
+  
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#extension GL_GOOGLE_include_directive : enable
+
+precision highp float;
+
+#include "common.h"
+
+layout(location = 0) in vec2 pos;
+layout(location = 1) in vec2 uv;
+
+layout(push_constant) uniform Block {
+	vec4 uvTransform;
+	vec4 subUVTransform;
+} _block;
+
+layout(location = 0) out vec2 vUV;
+
+void main() {
+  vec3 p = transform3x2(fboTransform, vec3(pos, 1.0));
+  gl_Position = vec4(p, 1);
+  vUV = uv.xy*_block.subUVTransform.xy + _block.subUVTransform.zw;
+  vUV = vUV*_block.uvTransform.xy + _block.uvTransform.zw;
+}

gio/shader/gio/material.frag

@@ -0,0 +1,32 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+precision mediump float;
+
+layout(binding = 0) uniform sampler2D tex;
+
+layout(location = 0) in highp vec2 vUV;
+
+layout(location = 0) out vec4 fragColor;
+
+layout(push_constant) uniform Color {
+	// If emulateSRGB is set (!= 0), the input texels are sRGB encoded. We save the
+	// conversion step below, at the cost of texture filtering in sRGB space.
+	layout(offset=16) float emulateSRGB;
+} _color;
+
+vec3 RGBtosRGB(vec3 rgb) {
+	bvec3 cutoff = greaterThanEqual(rgb, vec3(0.0031308));
+	vec3 below = vec3(12.92)*rgb;
+	vec3 above = vec3(1.055)*pow(rgb, vec3(0.41666)) - vec3(0.055);
+	return mix(below, above, cutoff);
+}
+
+void main() {
+	vec4 texel = texture(tex, vUV);
+	if (_color.emulateSRGB == 0.0) {
+		texel.rgb = RGBtosRGB(texel.rgb);
+	}
+	fragColor = texel;
+}

gio/shader/gio/material.vert

@@ -0,0 +1,25 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#extension GL_GOOGLE_include_directive : enable
+
+precision highp float;
+
+#include "common.h"
+
+layout(push_constant) uniform Block {
+	vec2 scale;
+	vec2 pos;
+} _block;
+
+layout(location = 0) in vec2 pos;
+layout(location = 1) in vec2 uv;
+
+layout(location = 0) out vec2 vUV;
+
+void main() {
+	vUV = uv;
+	vec2 p = vec2(pos*_block.scale + _block.pos);
+	gl_Position = vec4(transform3x2(fboTransform, vec3(p, 0)), 1);
+}

gio/shader/gio/shaders.go

@@ -0,0 +1,796 @@
+// Code generated by build.go. DO NOT EDIT.
+
+package gio
+
+import (
+	_ "embed"
+	"runtime"
+
+	"gioui.org/shader"
+)
+
+var (
+	Shader_blit_frag = [...]shader.Sources{
+		{
+			Name:   "blit.frag",
+			Inputs: []shader.InputLocation{{Name: "vUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}},
+			Uniforms: shader.UniformsReflection{
+				Locations: []shader.UniformLocation{{Name: "_color.color", Type: 0x0, Size: 4, Offset: 112}},
+				Size:      16,
+			},
+		},
+		{
+			Name:   "blit.frag",
+			Inputs: []shader.InputLocation{{Name: "vUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}},
+			Uniforms: shader.UniformsReflection{
+				Locations: []shader.UniformLocation{{Name: "_gradient.color1", Type: 0x0, Size: 4, Offset: 96}, {Name: "_gradient.color2", Type: 0x0, Size: 4, Offset: 112}},
+				Size:      32,
+			},
+		},
+		{
+			Name:     "blit.frag",
+			Inputs:   []shader.InputLocation{{Name: "vUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}},
+			Textures: []shader.TextureBinding{{Name: "tex", Binding: 0}},
+		},
+	}
+	//go:embed zblit.frag.0.spirv
+	zblit_frag_0_spirv string
+	//go:embed zblit.frag.0.glsl100es
+	zblit_frag_0_glsl100es string
+	//go:embed zblit.frag.0.glsl150
+	zblit_frag_0_glsl150 string
+	//go:embed zblit.frag.0.dxbc
+	zblit_frag_0_dxbc string
+	//go:embed zblit.frag.0.metallibmacos
+	zblit_frag_0_metallibmacos string
+	//go:embed zblit.frag.0.metallibios
+	zblit_frag_0_metallibios string
+	//go:embed zblit.frag.0.metallibiossimulator
+	zblit_frag_0_metallibiossimulator string
+	//go:embed zblit.frag.1.spirv
+	zblit_frag_1_spirv string
+	//go:embed zblit.frag.1.glsl100es
+	zblit_frag_1_glsl100es string
+	//go:embed zblit.frag.1.glsl150
+	zblit_frag_1_glsl150 string
+	//go:embed zblit.frag.1.dxbc
+	zblit_frag_1_dxbc string
+	//go:embed zblit.frag.1.metallibmacos
+	zblit_frag_1_metallibmacos string
+	//go:embed zblit.frag.1.metallibios
+	zblit_frag_1_metallibios string
+	//go:embed zblit.frag.1.metallibiossimulator
+	zblit_frag_1_metallibiossimulator string
+	//go:embed zblit.frag.2.spirv
+	zblit_frag_2_spirv string
+	//go:embed zblit.frag.2.glsl100es
+	zblit_frag_2_glsl100es string
+	//go:embed zblit.frag.2.glsl150
+	zblit_frag_2_glsl150 string
+	//go:embed zblit.frag.2.dxbc
+	zblit_frag_2_dxbc string
+	//go:embed zblit.frag.2.metallibmacos
+	zblit_frag_2_metallibmacos string
+	//go:embed zblit.frag.2.metallibios
+	zblit_frag_2_metallibios string
+	//go:embed zblit.frag.2.metallibiossimulator
+	zblit_frag_2_metallibiossimulator string
+	Shader_blit_vert                  = shader.Sources{
+		Name:   "blit.vert",
+		Inputs: []shader.InputLocation{{Name: "pos", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "uv", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+		Uniforms: shader.UniformsReflection{
+			Locations: []shader.UniformLocation{{Name: "_block.transform", Type: 0x0, Size: 4, Offset: 0}, {Name: "_block.uvTransformR1", Type: 0x0, Size: 4, Offset: 16}, {Name: "_block.uvTransformR2", Type: 0x0, Size: 4, Offset: 32}},
+			Size:      48,
+		},
+	}
+	//go:embed zblit.vert.0.spirv
+	zblit_vert_0_spirv string
+	//go:embed zblit.vert.0.glsl100es
+	zblit_vert_0_glsl100es string
+	//go:embed zblit.vert.0.glsl150
+	zblit_vert_0_glsl150 string
+	//go:embed zblit.vert.0.dxbc
+	zblit_vert_0_dxbc string
+	//go:embed zblit.vert.0.metallibmacos
+	zblit_vert_0_metallibmacos string
+	//go:embed zblit.vert.0.metallibios
+	zblit_vert_0_metallibios string
+	//go:embed zblit.vert.0.metallibiossimulator
+	zblit_vert_0_metallibiossimulator string
+	Shader_copy_frag                  = shader.Sources{
+		Name:     "copy.frag",
+		Inputs:   []shader.InputLocation{{Name: "vUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}},
+		Textures: []shader.TextureBinding{{Name: "tex", Binding: 0}},
+	}
+	//go:embed zcopy.frag.0.spirv
+	zcopy_frag_0_spirv string
+	//go:embed zcopy.frag.0.glsl100es
+	zcopy_frag_0_glsl100es string
+	//go:embed zcopy.frag.0.glsl150
+	zcopy_frag_0_glsl150 string
+	//go:embed zcopy.frag.0.dxbc
+	zcopy_frag_0_dxbc string
+	//go:embed zcopy.frag.0.metallibmacos
+	zcopy_frag_0_metallibmacos string
+	//go:embed zcopy.frag.0.metallibios
+	zcopy_frag_0_metallibios string
+	//go:embed zcopy.frag.0.metallibiossimulator
+	zcopy_frag_0_metallibiossimulator string
+	Shader_copy_vert                  = shader.Sources{
+		Name:   "copy.vert",
+		Inputs: []shader.InputLocation{{Name: "pos", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "uv", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+		Uniforms: shader.UniformsReflection{
+			Locations: []shader.UniformLocation{{Name: "_block.scale", Type: 0x0, Size: 2, Offset: 0}, {Name: "_block.pos", Type: 0x0, Size: 2, Offset: 8}, {Name: "_block.uvScale", Type: 0x0, Size: 2, Offset: 16}},
+			Size:      24,
+		},
+	}
+	//go:embed zcopy.vert.0.spirv
+	zcopy_vert_0_spirv string
+	//go:embed zcopy.vert.0.glsl100es
+	zcopy_vert_0_glsl100es string
+	//go:embed zcopy.vert.0.glsl150
+	zcopy_vert_0_glsl150 string
+	//go:embed zcopy.vert.0.dxbc
+	zcopy_vert_0_dxbc string
+	//go:embed zcopy.vert.0.metallibmacos
+	zcopy_vert_0_metallibmacos string
+	//go:embed zcopy.vert.0.metallibios
+	zcopy_vert_0_metallibios string
+	//go:embed zcopy.vert.0.metallibiossimulator
+	zcopy_vert_0_metallibiossimulator string
+	Shader_cover_frag                 = [...]shader.Sources{
+		{
+			Name:   "cover.frag",
+			Inputs: []shader.InputLocation{{Name: "vCoverUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "vUV", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+			Uniforms: shader.UniformsReflection{
+				Locations: []shader.UniformLocation{{Name: "_color.color", Type: 0x0, Size: 4, Offset: 112}},
+				Size:      16,
+			},
+			Textures: []shader.TextureBinding{{Name: "cover", Binding: 1}},
+		},
+		{
+			Name:   "cover.frag",
+			Inputs: []shader.InputLocation{{Name: "vCoverUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "vUV", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+			Uniforms: shader.UniformsReflection{
+				Locations: []shader.UniformLocation{{Name: "_gradient.color1", Type: 0x0, Size: 4, Offset: 96}, {Name: "_gradient.color2", Type: 0x0, Size: 4, Offset: 112}},
+				Size:      32,
+			},
+			Textures: []shader.TextureBinding{{Name: "cover", Binding: 1}},
+		},
+		{
+			Name:     "cover.frag",
+			Inputs:   []shader.InputLocation{{Name: "vCoverUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "vUV", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+			Textures: []shader.TextureBinding{{Name: "tex", Binding: 0}, {Name: "cover", Binding: 1}},
+		},
+	}
+	//go:embed zcover.frag.0.spirv
+	zcover_frag_0_spirv string
+	//go:embed zcover.frag.0.glsl100es
+	zcover_frag_0_glsl100es string
+	//go:embed zcover.frag.0.glsl150
+	zcover_frag_0_glsl150 string
+	//go:embed zcover.frag.0.dxbc
+	zcover_frag_0_dxbc string
+	//go:embed zcover.frag.0.metallibmacos
+	zcover_frag_0_metallibmacos string
+	//go:embed zcover.frag.0.metallibios
+	zcover_frag_0_metallibios string
+	//go:embed zcover.frag.0.metallibiossimulator
+	zcover_frag_0_metallibiossimulator string
+	//go:embed zcover.frag.1.spirv
+	zcover_frag_1_spirv string
+	//go:embed zcover.frag.1.glsl100es
+	zcover_frag_1_glsl100es string
+	//go:embed zcover.frag.1.glsl150
+	zcover_frag_1_glsl150 string
+	//go:embed zcover.frag.1.dxbc
+	zcover_frag_1_dxbc string
+	//go:embed zcover.frag.1.metallibmacos
+	zcover_frag_1_metallibmacos string
+	//go:embed zcover.frag.1.metallibios
+	zcover_frag_1_metallibios string
+	//go:embed zcover.frag.1.metallibiossimulator
+	zcover_frag_1_metallibiossimulator string
+	//go:embed zcover.frag.2.spirv
+	zcover_frag_2_spirv string
+	//go:embed zcover.frag.2.glsl100es
+	zcover_frag_2_glsl100es string
+	//go:embed zcover.frag.2.glsl150
+	zcover_frag_2_glsl150 string
+	//go:embed zcover.frag.2.dxbc
+	zcover_frag_2_dxbc string
+	//go:embed zcover.frag.2.metallibmacos
+	zcover_frag_2_metallibmacos string
+	//go:embed zcover.frag.2.metallibios
+	zcover_frag_2_metallibios string
+	//go:embed zcover.frag.2.metallibiossimulator
+	zcover_frag_2_metallibiossimulator string
+	Shader_cover_vert                  = shader.Sources{
+		Name:   "cover.vert",
+		Inputs: []shader.InputLocation{{Name: "pos", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "uv", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+		Uniforms: shader.UniformsReflection{
+			Locations: []shader.UniformLocation{{Name: "_block.transform", Type: 0x0, Size: 4, Offset: 0}, {Name: "_block.uvCoverTransform", Type: 0x0, Size: 4, Offset: 16}, {Name: "_block.uvTransformR1", Type: 0x0, Size: 4, Offset: 32}, {Name: "_block.uvTransformR2", Type: 0x0, Size: 4, Offset: 48}},
+			Size:      64,
+		},
+	}
+	//go:embed zcover.vert.0.spirv
+	zcover_vert_0_spirv string
+	//go:embed zcover.vert.0.glsl100es
+	zcover_vert_0_glsl100es string
+	//go:embed zcover.vert.0.glsl150
+	zcover_vert_0_glsl150 string
+	//go:embed zcover.vert.0.dxbc
+	zcover_vert_0_dxbc string
+	//go:embed zcover.vert.0.metallibmacos
+	zcover_vert_0_metallibmacos string
+	//go:embed zcover.vert.0.metallibios
+	zcover_vert_0_metallibios string
+	//go:embed zcover.vert.0.metallibiossimulator
+	zcover_vert_0_metallibiossimulator string
+	Shader_input_vert                  = shader.Sources{
+		Name:   "input.vert",
+		Inputs: []shader.InputLocation{{Name: "position", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 4}},
+	}
+	//go:embed zinput.vert.0.spirv
+	zinput_vert_0_spirv string
+	//go:embed zinput.vert.0.glsl100es
+	zinput_vert_0_glsl100es string
+	//go:embed zinput.vert.0.glsl150
+	zinput_vert_0_glsl150 string
+	//go:embed zinput.vert.0.dxbc
+	zinput_vert_0_dxbc string
+	//go:embed zinput.vert.0.metallibmacos
+	zinput_vert_0_metallibmacos string
+	//go:embed zinput.vert.0.metallibios
+	zinput_vert_0_metallibios string
+	//go:embed zinput.vert.0.metallibiossimulator
+	zinput_vert_0_metallibiossimulator string
+	Shader_intersect_frag              = shader.Sources{
+		Name:     "intersect.frag",
+		Inputs:   []shader.InputLocation{{Name: "vUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}},
+		Textures: []shader.TextureBinding{{Name: "cover", Binding: 0}},
+	}
+	//go:embed zintersect.frag.0.spirv
+	zintersect_frag_0_spirv string
+	//go:embed zintersect.frag.0.glsl100es
+	zintersect_frag_0_glsl100es string
+	//go:embed zintersect.frag.0.glsl150
+	zintersect_frag_0_glsl150 string
+	//go:embed zintersect.frag.0.dxbc
+	zintersect_frag_0_dxbc string
+	//go:embed zintersect.frag.0.metallibmacos
+	zintersect_frag_0_metallibmacos string
+	//go:embed zintersect.frag.0.metallibios
+	zintersect_frag_0_metallibios string
+	//go:embed zintersect.frag.0.metallibiossimulator
+	zintersect_frag_0_metallibiossimulator string
+	Shader_intersect_vert                  = shader.Sources{
+		Name:   "intersect.vert",
+		Inputs: []shader.InputLocation{{Name: "pos", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "uv", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+		Uniforms: shader.UniformsReflection{
+			Locations: []shader.UniformLocation{{Name: "_block.uvTransform", Type: 0x0, Size: 4, Offset: 0}, {Name: "_block.subUVTransform", Type: 0x0, Size: 4, Offset: 16}},
+			Size:      32,
+		},
+	}
+	//go:embed zintersect.vert.0.spirv
+	zintersect_vert_0_spirv string
+	//go:embed zintersect.vert.0.glsl100es
+	zintersect_vert_0_glsl100es string
+	//go:embed zintersect.vert.0.glsl150
+	zintersect_vert_0_glsl150 string
+	//go:embed zintersect.vert.0.dxbc
+	zintersect_vert_0_dxbc string
+	//go:embed zintersect.vert.0.metallibmacos
+	zintersect_vert_0_metallibmacos string
+	//go:embed zintersect.vert.0.metallibios
+	zintersect_vert_0_metallibios string
+	//go:embed zintersect.vert.0.metallibiossimulator
+	zintersect_vert_0_metallibiossimulator string
+	Shader_material_frag                   = shader.Sources{
+		Name:   "material.frag",
+		Inputs: []shader.InputLocation{{Name: "vUV", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}},
+		Uniforms: shader.UniformsReflection{
+			Locations: []shader.UniformLocation{{Name: "_color.emulateSRGB", Type: 0x0, Size: 1, Offset: 16}},
+			Size:      4,
+		},
+		Textures: []shader.TextureBinding{{Name: "tex", Binding: 0}},
+	}
+	//go:embed zmaterial.frag.0.spirv
+	zmaterial_frag_0_spirv string
+	//go:embed zmaterial.frag.0.glsl100es
+	zmaterial_frag_0_glsl100es string
+	//go:embed zmaterial.frag.0.glsl150
+	zmaterial_frag_0_glsl150 string
+	//go:embed zmaterial.frag.0.dxbc
+	zmaterial_frag_0_dxbc string
+	//go:embed zmaterial.frag.0.metallibmacos
+	zmaterial_frag_0_metallibmacos string
+	//go:embed zmaterial.frag.0.metallibios
+	zmaterial_frag_0_metallibios string
+	//go:embed zmaterial.frag.0.metallibiossimulator
+	zmaterial_frag_0_metallibiossimulator string
+	Shader_material_vert                  = shader.Sources{
+		Name:   "material.vert",
+		Inputs: []shader.InputLocation{{Name: "pos", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "uv", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}},
+		Uniforms: shader.UniformsReflection{
+			Locations: []shader.UniformLocation{{Name: "_block.scale", Type: 0x0, Size: 2, Offset: 0}, {Name: "_block.pos", Type: 0x0, Size: 2, Offset: 8}},
+			Size:      16,
+		},
+	}
+	//go:embed zmaterial.vert.0.spirv
+	zmaterial_vert_0_spirv string
+	//go:embed zmaterial.vert.0.glsl100es
+	zmaterial_vert_0_glsl100es string
+	//go:embed zmaterial.vert.0.glsl150
+	zmaterial_vert_0_glsl150 string
+	//go:embed zmaterial.vert.0.dxbc
+	zmaterial_vert_0_dxbc string
+	//go:embed zmaterial.vert.0.metallibmacos
+	zmaterial_vert_0_metallibmacos string
+	//go:embed zmaterial.vert.0.metallibios
+	zmaterial_vert_0_metallibios string
+	//go:embed zmaterial.vert.0.metallibiossimulator
+	zmaterial_vert_0_metallibiossimulator string
+	Shader_simple_frag                    = shader.Sources{
+		Name: "simple.frag",
+	}
+	//go:embed zsimple.frag.0.spirv
+	zsimple_frag_0_spirv string
+	//go:embed zsimple.frag.0.glsl100es
+	zsimple_frag_0_glsl100es string
+	//go:embed zsimple.frag.0.glsl150
+	zsimple_frag_0_glsl150 string
+	//go:embed zsimple.frag.0.dxbc
+	zsimple_frag_0_dxbc string
+	//go:embed zsimple.frag.0.metallibmacos
+	zsimple_frag_0_metallibmacos string
+	//go:embed zsimple.frag.0.metallibios
+	zsimple_frag_0_metallibios string
+	//go:embed zsimple.frag.0.metallibiossimulator
+	zsimple_frag_0_metallibiossimulator string
+	Shader_stencil_frag                 = shader.Sources{
+		Name:   "stencil.frag",
+		Inputs: []shader.InputLocation{{Name: "vFrom", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 2}, {Name: "vCtrl", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 2}, {Name: "vTo", Location: 2, Semantic: "TEXCOORD", SemanticIndex: 2, Type: 0x0, Size: 2}},
+	}
+	//go:embed zstencil.frag.0.spirv
+	zstencil_frag_0_spirv string
+	//go:embed zstencil.frag.0.glsl100es
+	zstencil_frag_0_glsl100es string
+	//go:embed zstencil.frag.0.glsl150
+	zstencil_frag_0_glsl150 string
+	//go:embed zstencil.frag.0.dxbc
+	zstencil_frag_0_dxbc string
+	//go:embed zstencil.frag.0.metallibmacos
+	zstencil_frag_0_metallibmacos string
+	//go:embed zstencil.frag.0.metallibios
+	zstencil_frag_0_metallibios string
+	//go:embed zstencil.frag.0.metallibiossimulator
+	zstencil_frag_0_metallibiossimulator string
+	Shader_stencil_vert                  = shader.Sources{
+		Name:   "stencil.vert",
+		Inputs: []shader.InputLocation{{Name: "corner", Location: 0, Semantic: "TEXCOORD", SemanticIndex: 0, Type: 0x0, Size: 1}, {Name: "maxy", Location: 1, Semantic: "TEXCOORD", SemanticIndex: 1, Type: 0x0, Size: 1}, {Name: "from", Location: 2, Semantic: "TEXCOORD", SemanticIndex: 2, Type: 0x0, Size: 2}, {Name: "ctrl", Location: 3, Semantic: "TEXCOORD", SemanticIndex: 3, Type: 0x0, Size: 2}, {Name: "to", Location: 4, Semantic: "TEXCOORD", SemanticIndex: 4, Type: 0x0, Size: 2}},
+		Uniforms: shader.UniformsReflection{
+			Locations: []shader.UniformLocation{{Name: "_block.transform", Type: 0x0, Size: 4, Offset: 0}, {Name: "_block.pathOffset", Type: 0x0, Size: 2, Offset: 16}},
+			Size:      24,
+		},
+	}
+	//go:embed zstencil.vert.0.spirv
+	zstencil_vert_0_spirv string
+	//go:embed zstencil.vert.0.glsl100es
+	zstencil_vert_0_glsl100es string
+	//go:embed zstencil.vert.0.glsl150
+	zstencil_vert_0_glsl150 string
+	//go:embed zstencil.vert.0.dxbc
+	zstencil_vert_0_dxbc string
+	//go:embed zstencil.vert.0.metallibmacos
+	zstencil_vert_0_metallibmacos string
+	//go:embed zstencil.vert.0.metallibios
+	zstencil_vert_0_metallibios string
+	//go:embed zstencil.vert.0.metallibiossimulator
+	zstencil_vert_0_metallibiossimulator string
+)
+
+func init() {
+	const (
+		opengles = runtime.GOOS == "linux" || runtime.GOOS == "freebsd" || runtime.GOOS == "openbsd" || runtime.GOOS == "windows" || runtime.GOOS == "js" || runtime.GOOS == "android" || runtime.GOOS == "darwin" || runtime.GOOS == "ios"
+		opengl   = runtime.GOOS == "darwin"
+		d3d11    = runtime.GOOS == "windows"
+		vulkan   = runtime.GOOS == "linux" || runtime.GOOS == "android"
+	)
+	if vulkan {
+		Shader_blit_frag[0].SPIRV = zblit_frag_0_spirv
+	}
+	if opengles {
+		Shader_blit_frag[0].GLSL100ES = zblit_frag_0_glsl100es
+	}
+	if opengl {
+		Shader_blit_frag[0].GLSL150 = zblit_frag_0_glsl150
+	}
+	if d3d11 {
+		Shader_blit_frag[0].DXBC = zblit_frag_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_blit_frag[0].MetalLib = zblit_frag_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_blit_frag[0].MetalLib = zblit_frag_0_metallibiossimulator
+		} else {
+			Shader_blit_frag[0].MetalLib = zblit_frag_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_blit_frag[1].SPIRV = zblit_frag_1_spirv
+	}
+	if opengles {
+		Shader_blit_frag[1].GLSL100ES = zblit_frag_1_glsl100es
+	}
+	if opengl {
+		Shader_blit_frag[1].GLSL150 = zblit_frag_1_glsl150
+	}
+	if d3d11 {
+		Shader_blit_frag[1].DXBC = zblit_frag_1_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_blit_frag[1].MetalLib = zblit_frag_1_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_blit_frag[1].MetalLib = zblit_frag_1_metallibiossimulator
+		} else {
+			Shader_blit_frag[1].MetalLib = zblit_frag_1_metallibios
+		}
+	}
+	if vulkan {
+		Shader_blit_frag[2].SPIRV = zblit_frag_2_spirv
+	}
+	if opengles {
+		Shader_blit_frag[2].GLSL100ES = zblit_frag_2_glsl100es
+	}
+	if opengl {
+		Shader_blit_frag[2].GLSL150 = zblit_frag_2_glsl150
+	}
+	if d3d11 {
+		Shader_blit_frag[2].DXBC = zblit_frag_2_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_blit_frag[2].MetalLib = zblit_frag_2_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_blit_frag[2].MetalLib = zblit_frag_2_metallibiossimulator
+		} else {
+			Shader_blit_frag[2].MetalLib = zblit_frag_2_metallibios
+		}
+	}
+	if vulkan {
+		Shader_blit_vert.SPIRV = zblit_vert_0_spirv
+	}
+	if opengles {
+		Shader_blit_vert.GLSL100ES = zblit_vert_0_glsl100es
+	}
+	if opengl {
+		Shader_blit_vert.GLSL150 = zblit_vert_0_glsl150
+	}
+	if d3d11 {
+		Shader_blit_vert.DXBC = zblit_vert_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_blit_vert.MetalLib = zblit_vert_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_blit_vert.MetalLib = zblit_vert_0_metallibiossimulator
+		} else {
+			Shader_blit_vert.MetalLib = zblit_vert_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_copy_frag.SPIRV = zcopy_frag_0_spirv
+	}
+	if opengles {
+		Shader_copy_frag.GLSL100ES = zcopy_frag_0_glsl100es
+	}
+	if opengl {
+		Shader_copy_frag.GLSL150 = zcopy_frag_0_glsl150
+	}
+	if d3d11 {
+		Shader_copy_frag.DXBC = zcopy_frag_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_copy_frag.MetalLib = zcopy_frag_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_copy_frag.MetalLib = zcopy_frag_0_metallibiossimulator
+		} else {
+			Shader_copy_frag.MetalLib = zcopy_frag_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_copy_vert.SPIRV = zcopy_vert_0_spirv
+	}
+	if opengles {
+		Shader_copy_vert.GLSL100ES = zcopy_vert_0_glsl100es
+	}
+	if opengl {
+		Shader_copy_vert.GLSL150 = zcopy_vert_0_glsl150
+	}
+	if d3d11 {
+		Shader_copy_vert.DXBC = zcopy_vert_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_copy_vert.MetalLib = zcopy_vert_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_copy_vert.MetalLib = zcopy_vert_0_metallibiossimulator
+		} else {
+			Shader_copy_vert.MetalLib = zcopy_vert_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_cover_frag[0].SPIRV = zcover_frag_0_spirv
+	}
+	if opengles {
+		Shader_cover_frag[0].GLSL100ES = zcover_frag_0_glsl100es
+	}
+	if opengl {
+		Shader_cover_frag[0].GLSL150 = zcover_frag_0_glsl150
+	}
+	if d3d11 {
+		Shader_cover_frag[0].DXBC = zcover_frag_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_cover_frag[0].MetalLib = zcover_frag_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_cover_frag[0].MetalLib = zcover_frag_0_metallibiossimulator
+		} else {
+			Shader_cover_frag[0].MetalLib = zcover_frag_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_cover_frag[1].SPIRV = zcover_frag_1_spirv
+	}
+	if opengles {
+		Shader_cover_frag[1].GLSL100ES = zcover_frag_1_glsl100es
+	}
+	if opengl {
+		Shader_cover_frag[1].GLSL150 = zcover_frag_1_glsl150
+	}
+	if d3d11 {
+		Shader_cover_frag[1].DXBC = zcover_frag_1_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_cover_frag[1].MetalLib = zcover_frag_1_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_cover_frag[1].MetalLib = zcover_frag_1_metallibiossimulator
+		} else {
+			Shader_cover_frag[1].MetalLib = zcover_frag_1_metallibios
+		}
+	}
+	if vulkan {
+		Shader_cover_frag[2].SPIRV = zcover_frag_2_spirv
+	}
+	if opengles {
+		Shader_cover_frag[2].GLSL100ES = zcover_frag_2_glsl100es
+	}
+	if opengl {
+		Shader_cover_frag[2].GLSL150 = zcover_frag_2_glsl150
+	}
+	if d3d11 {
+		Shader_cover_frag[2].DXBC = zcover_frag_2_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_cover_frag[2].MetalLib = zcover_frag_2_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_cover_frag[2].MetalLib = zcover_frag_2_metallibiossimulator
+		} else {
+			Shader_cover_frag[2].MetalLib = zcover_frag_2_metallibios
+		}
+	}
+	if vulkan {
+		Shader_cover_vert.SPIRV = zcover_vert_0_spirv
+	}
+	if opengles {
+		Shader_cover_vert.GLSL100ES = zcover_vert_0_glsl100es
+	}
+	if opengl {
+		Shader_cover_vert.GLSL150 = zcover_vert_0_glsl150
+	}
+	if d3d11 {
+		Shader_cover_vert.DXBC = zcover_vert_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_cover_vert.MetalLib = zcover_vert_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_cover_vert.MetalLib = zcover_vert_0_metallibiossimulator
+		} else {
+			Shader_cover_vert.MetalLib = zcover_vert_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_input_vert.SPIRV = zinput_vert_0_spirv
+	}
+	if opengles {
+		Shader_input_vert.GLSL100ES = zinput_vert_0_glsl100es
+	}
+	if opengl {
+		Shader_input_vert.GLSL150 = zinput_vert_0_glsl150
+	}
+	if d3d11 {
+		Shader_input_vert.DXBC = zinput_vert_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_input_vert.MetalLib = zinput_vert_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_input_vert.MetalLib = zinput_vert_0_metallibiossimulator
+		} else {
+			Shader_input_vert.MetalLib = zinput_vert_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_intersect_frag.SPIRV = zintersect_frag_0_spirv
+	}
+	if opengles {
+		Shader_intersect_frag.GLSL100ES = zintersect_frag_0_glsl100es
+	}
+	if opengl {
+		Shader_intersect_frag.GLSL150 = zintersect_frag_0_glsl150
+	}
+	if d3d11 {
+		Shader_intersect_frag.DXBC = zintersect_frag_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_intersect_frag.MetalLib = zintersect_frag_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_intersect_frag.MetalLib = zintersect_frag_0_metallibiossimulator
+		} else {
+			Shader_intersect_frag.MetalLib = zintersect_frag_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_intersect_vert.SPIRV = zintersect_vert_0_spirv
+	}
+	if opengles {
+		Shader_intersect_vert.GLSL100ES = zintersect_vert_0_glsl100es
+	}
+	if opengl {
+		Shader_intersect_vert.GLSL150 = zintersect_vert_0_glsl150
+	}
+	if d3d11 {
+		Shader_intersect_vert.DXBC = zintersect_vert_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_intersect_vert.MetalLib = zintersect_vert_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_intersect_vert.MetalLib = zintersect_vert_0_metallibiossimulator
+		} else {
+			Shader_intersect_vert.MetalLib = zintersect_vert_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_material_frag.SPIRV = zmaterial_frag_0_spirv
+	}
+	if opengles {
+		Shader_material_frag.GLSL100ES = zmaterial_frag_0_glsl100es
+	}
+	if opengl {
+		Shader_material_frag.GLSL150 = zmaterial_frag_0_glsl150
+	}
+	if d3d11 {
+		Shader_material_frag.DXBC = zmaterial_frag_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_material_frag.MetalLib = zmaterial_frag_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_material_frag.MetalLib = zmaterial_frag_0_metallibiossimulator
+		} else {
+			Shader_material_frag.MetalLib = zmaterial_frag_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_material_vert.SPIRV = zmaterial_vert_0_spirv
+	}
+	if opengles {
+		Shader_material_vert.GLSL100ES = zmaterial_vert_0_glsl100es
+	}
+	if opengl {
+		Shader_material_vert.GLSL150 = zmaterial_vert_0_glsl150
+	}
+	if d3d11 {
+		Shader_material_vert.DXBC = zmaterial_vert_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_material_vert.MetalLib = zmaterial_vert_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_material_vert.MetalLib = zmaterial_vert_0_metallibiossimulator
+		} else {
+			Shader_material_vert.MetalLib = zmaterial_vert_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_simple_frag.SPIRV = zsimple_frag_0_spirv
+	}
+	if opengles {
+		Shader_simple_frag.GLSL100ES = zsimple_frag_0_glsl100es
+	}
+	if opengl {
+		Shader_simple_frag.GLSL150 = zsimple_frag_0_glsl150
+	}
+	if d3d11 {
+		Shader_simple_frag.DXBC = zsimple_frag_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_simple_frag.MetalLib = zsimple_frag_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_simple_frag.MetalLib = zsimple_frag_0_metallibiossimulator
+		} else {
+			Shader_simple_frag.MetalLib = zsimple_frag_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_stencil_frag.SPIRV = zstencil_frag_0_spirv
+	}
+	if opengles {
+		Shader_stencil_frag.GLSL100ES = zstencil_frag_0_glsl100es
+	}
+	if opengl {
+		Shader_stencil_frag.GLSL150 = zstencil_frag_0_glsl150
+	}
+	if d3d11 {
+		Shader_stencil_frag.DXBC = zstencil_frag_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_stencil_frag.MetalLib = zstencil_frag_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_stencil_frag.MetalLib = zstencil_frag_0_metallibiossimulator
+		} else {
+			Shader_stencil_frag.MetalLib = zstencil_frag_0_metallibios
+		}
+	}
+	if vulkan {
+		Shader_stencil_vert.SPIRV = zstencil_vert_0_spirv
+	}
+	if opengles {
+		Shader_stencil_vert.GLSL100ES = zstencil_vert_0_glsl100es
+	}
+	if opengl {
+		Shader_stencil_vert.GLSL150 = zstencil_vert_0_glsl150
+	}
+	if d3d11 {
+		Shader_stencil_vert.DXBC = zstencil_vert_0_dxbc
+	}
+	if runtime.GOOS == "darwin" {
+		Shader_stencil_vert.MetalLib = zstencil_vert_0_metallibmacos
+	}
+	if runtime.GOOS == "ios" {
+		if runtime.GOARCH == "amd64" {
+			Shader_stencil_vert.MetalLib = zstencil_vert_0_metallibiossimulator
+		} else {
+			Shader_stencil_vert.MetalLib = zstencil_vert_0_metallibios
+		}
+	}
+}

gio/shader/gio/simple.frag

@@ -0,0 +1,11 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+precision mediump float;
+
+layout(location = 0) out vec4 fragColor;
+
+void main() {
+	fragColor = vec4(.25, .55, .75, 1.0);
+}

gio/shader/gio/stencil.frag

@@ -0,0 +1,81 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+precision mediump float;
+
+layout(location=0) in highp vec2 vFrom;
+layout(location=1) in highp vec2 vCtrl;
+layout(location=2) in highp vec2 vTo;
+
+layout(location = 0) out vec4 fragCover;
+
+void main() {
+	float dx = vTo.x - vFrom.x;
+	// Sort from and to in increasing order so the root below
+	// is always the positive square root, if any.
+	// We need the direction of the curve below, so this can't be
+	// done from the vertex shader.
+	bool increasing = vTo.x >= vFrom.x;
+	vec2 left = increasing ? vFrom : vTo;
+	vec2 right = increasing ? vTo : vFrom;
+
+	// The signed horizontal extent of the fragment.
+	vec2 extent = clamp(vec2(vFrom.x, vTo.x), -0.5, 0.5);
+	// Find the t where the curve crosses the middle of the
+	// extent, x₀.
+	// Given the Bézier curve with x coordinates P₀, P₁, P₂
+	// where P₀ is at the origin, its x coordinate in t
+	// is given by:
+	//
+	// x(t) = 2(1-t)tP₁ + t²P₂
+	// 
+	// Rearranging:
+	//
+	// x(t) = (P₂ - 2P₁)t² + 2P₁t
+	//
+	// Setting x(t) = x₀ and using Muller's quadratic formula ("Citardauq")
+	// for robustnesss,
+	//
+	// t = 2x₀/(2P₁±√(4P₁²+4(P₂-2P₁)x₀))
+	//
+	// which simplifies to
+	//
+	// t = x₀/(P₁±√(P₁²+(P₂-2P₁)x₀))
+	//
+	// Setting v = P₂-P₁,
+	//
+	// t = x₀/(P₁±√(P₁²+(v-P₁)x₀))
+	//
+	// t lie in [0; 1]; P₂ ≥ P₁ and P₁ ≥ 0 since we split curves where
+	// the control point lies before the start point or after the end point.
+	// It can then be shown that only the positive square root is valid.
+	float midx = mix(extent.x, extent.y, 0.5);
+	float x0 = midx - left.x;
+	vec2 p1 = vCtrl - left;
+	vec2 v = right - vCtrl;
+	float t = x0/(p1.x+sqrt(p1.x*p1.x+(v.x-p1.x)*x0));
+	// Find y(t) on the curve.
+	float y = mix(mix(left.y, vCtrl.y, t), mix(vCtrl.y, right.y, t), t);
+	// And the slope.
+	vec2 d_half = mix(p1, v, t);
+	float dy = d_half.y/d_half.x;
+	// Together, y and dy form a line approximation.
+
+	// Compute the fragment area above the line.
+	// The area is symmetric around dy = 0. Scale slope with extent width.
+	float width = extent.y - extent.x;
+	dy = abs(dy*width);
+
+	vec4 sides = vec4(dy*+0.5 + y, dy*-0.5 + y, (+0.5-y)/dy, (-0.5-y)/dy);
+	sides = clamp(sides+0.5, 0.0, 1.0);
+
+	float area = 0.5*(sides.z - sides.z*sides.y + 1.0 - sides.x+sides.x*sides.w);
+	area *= width;
+
+	// Work around issue #13.
+	if (width == 0.0)
+		area = 0.0;
+
+	fragCover.r = area;
+}

gio/shader/gio/stencil.vert

@@ -0,0 +1,57 @@
+#version 310 es
+
+// SPDX-License-Identifier: Unlicense OR MIT
+
+#extension GL_GOOGLE_include_directive : enable
+
+precision highp float;
+
+#include "common.h"
+
+layout(push_constant) uniform Block {
+	vec4 transform;
+	vec2 pathOffset;
+} _block;
+
+layout(location=0) in float corner;
+layout(location=1) in float maxy;
+layout(location=2) in vec2 from;
+layout(location=3) in vec2 ctrl;
+layout(location=4) in vec2 to;
+
+layout(location=0) out vec2 vFrom;
+layout(location=1) out vec2 vCtrl;
+layout(location=2) out vec2 vTo;
+
+void main() {
+	// Add a one pixel overlap so curve quads cover their
+	// entire curves. Could use conservative rasterization
+	// if available.
+	vec2 from = from + _block.pathOffset;
+	vec2 ctrl = ctrl + _block.pathOffset;
+	vec2 to = to + _block.pathOffset;
+	float maxy = maxy + _block.pathOffset.y;
+	vec2 pos;
+	float c = corner;
+	if (c >= 0.375) {
+		// North.
+		c -= 0.5;
+		pos.y = maxy + 1.0;
+	} else {
+		// South.
+		pos.y = min(min(from.y, ctrl.y), to.y) - 1.0;
+	}
+	if (c >= 0.125) {
+		// East.
+		pos.x = max(max(from.x, ctrl.x), to.x)+1.0;
+	} else {
+		// West.
+		pos.x = min(min(from.x, ctrl.x), to.x)-1.0;
+	}
+	vFrom = from-pos;
+	vCtrl = ctrl-pos;
+	vTo = to-pos;
+	pos = pos*_block.transform.xy + _block.transform.zw;
+	gl_Position = vec4(transform3x2(fboTransform, vec3(pos, 0)), 1);
+}
+

gio/shader/gio/zblit.frag.0.glsl100es

@@ -0,0 +1,18 @@
+#version 100
+precision mediump float;
+precision highp int;
+
+struct Color
+{
+    vec4 color;
+};
+
+uniform Color _color;
+
+varying highp vec2 vUV;
+
+void main()
+{
+    gl_FragData[0] = _color.color;
+}
+

gio/shader/gio/zblit.frag.0.glsl150

@@ -0,0 +1,17 @@
+#version 150
+
+struct Color
+{
+    vec4 color;
+};
+
+uniform Color _color;
+
+out vec4 fragColor;
+in vec2 vUV;
+
+void main()
+{
+    fragColor = _color.color;
+}
+

gio/shader/gio/zblit.frag.1.glsl100es

@@ -0,0 +1,19 @@
+#version 100
+precision mediump float;
+precision highp int;
+
+struct Gradient
+{
+    vec4 color1;
+    vec4 color2;
+};
+
+uniform Gradient _gradient;
+
+varying highp vec2 vUV;
+
+void main()
+{
+    gl_FragData[0] = mix(_gradient.color1, _gradient.color2, vec4(clamp(vUV.x, 0.0, 1.0)));
+}
+

gio/shader/gio/zblit.frag.1.glsl150

@@ -0,0 +1,18 @@
+#version 150
+
+struct Gradient
+{
+    vec4 color1;
+    vec4 color2;
+};
+
+uniform Gradient _gradient;
+
+out vec4 fragColor;
+in vec2 vUV;
+
+void main()
+{
+    fragColor = mix(_gradient.color1, _gradient.color2, vec4(clamp(vUV.x, 0.0, 1.0)));
+}
+

gio/shader/gio/zblit.frag.2.glsl100es

@@ -0,0 +1,13 @@
+#version 100
+precision mediump float;
+precision highp int;
+
+uniform mediump sampler2D tex;
+
+varying highp vec2 vUV;
+
+void main()
+{
+    gl_FragData[0] = texture2D(tex, vUV);
+}
+

gio/shader/gio/zblit.frag.2.glsl150

@@ -0,0 +1,12 @@
+#version 150
+
+uniform sampler2D tex;
+
+out vec4 fragColor;
+in vec2 vUV;
+
+void main()
+{
+    fragColor = texture(tex, vUV);
+}
+

gio/shader/gio/zblit.vert.0.glsl100es

@@ -0,0 +1,37 @@
+#version 100
+
+struct m3x2
+{
+    vec3 r0;
+    vec3 r1;
+};
+
+struct Block
+{
+    vec4 transform;
+    vec4 uvTransformR1;
+    vec4 uvTransformR2;
+};
+
+uniform Block _block;
+
+attribute vec2 pos;
+varying vec2 vUV;
+attribute vec2 uv;
+
+vec3 transform3x2(m3x2 t, vec3 v)
+{
+    return vec3(dot(t.r0, v), dot(t.r1, v), dot(vec3(0.0, 0.0, 1.0), v));
+}
+
+void main()
+{
+    vec2 p = (pos * _block.transform.xy) + _block.transform.zw;
+    m3x2 param = m3x2(vec3(1.0, 0.0, 0.0), vec3(0.0, -1.0, 0.0));
+    vec3 param_1 = vec3(p, 0.0);
+    gl_Position = vec4(transform3x2(param, param_1), 1.0);
+    m3x2 param_2 = m3x2(_block.uvTransformR1.xyz, _block.uvTransformR2.xyz);
+    vec3 param_3 = vec3(uv, 1.0);
+    vUV = transform3x2(param_2, param_3).xy;
+}
+

gio/shader/gio/zblit.vert.0.glsl150

@@ -0,0 +1,37 @@
+#version 150
+
+struct m3x2
+{
+    vec3 r0;
+    vec3 r1;
+};
+
+struct Block
+{
+    vec4 transform;
+    vec4 uvTransformR1;
+    vec4 uvTransformR2;
+};
+
+uniform Block _block;
+
+in vec2 pos;
+out vec2 vUV;
+in vec2 uv;
+
+vec3 transform3x2(m3x2 t, vec3 v)
+{
+    return vec3(dot(t.r0, v), dot(t.r1, v), dot(vec3(0.0, 0.0, 1.0), v));
+}
+
+void main()
+{
+    vec2 p = (pos * _block.transform.xy) + _block.transform.zw;
+    m3x2 param = m3x2(vec3(1.0, 0.0, 0.0), vec3(0.0, -1.0, 0.0));
+    vec3 param_1 = vec3(p, 0.0);
+    gl_Position = vec4(transform3x2(param, param_1), 1.0);
+    m3x2 param_2 = m3x2(_block.uvTransformR1.xyz, _block.uvTransformR2.xyz);
+    vec3 param_3 = vec3(uv, 1.0);
+    vUV = transform3x2(param_2, param_3).xy;
+}
+

gio/shader/gio/zcopy.frag.0.glsl100es

@@ -0,0 +1,27 @@
+#version 100
+precision mediump float;
+precision highp int;
+
+uniform mediump sampler2D tex;
+
+varying highp vec2 vUV;
+
+vec3 sRGBtoRGB(vec3 rgb)
+{
+    bvec3 cutoff = greaterThanEqual(rgb, vec3(0.040449999272823333740234375));
+    vec3 below = rgb / vec3(12.9200000762939453125);
+    vec3 above = pow((rgb + vec3(0.054999999701976776123046875)) / vec3(1.05499994754791259765625), vec3(2.400000095367431640625));
+    return vec3(cutoff.x ? above.x : below.x, cutoff.y ? above.y : below.y, cutoff.z ? above.z : below.z);
+}
+
+void main()
+{
+    vec4 texel = texture2D(tex, vUV);
+    vec3 param = texel.xyz;
+    vec3 _59 = sRGBtoRGB(param);
+    texel.x = _59.x;
+    texel.y = _59.y;
+    texel.z = _59.z;
+    gl_FragData[0] = texel;
+}
+

gio/shader/gio/zcopy.frag.0.glsl150

@@ -0,0 +1,26 @@
+#version 150
+
+uniform sampler2D tex;
+
+in vec2 vUV;
+out vec4 fragColor;
+
+vec3 sRGBtoRGB(vec3 rgb)
+{
+    bvec3 cutoff = greaterThanEqual(rgb, vec3(0.040449999272823333740234375));
+    vec3 below = rgb / vec3(12.9200000762939453125);
+    vec3 above = pow((rgb + vec3(0.054999999701976776123046875)) / vec3(1.05499994754791259765625), vec3(2.400000095367431640625));
+    return vec3(cutoff.x ? above.x : below.x, cutoff.y ? above.y : below.y, cutoff.z ? above.z : below.z);
+}
+
+void main()
+{
+    vec4 texel = texture(tex, vUV);
+    vec3 param = texel.xyz;
+    vec3 _59 = sRGBtoRGB(param);
+    texel.x = _59.x;
+    texel.y = _59.y;
+    texel.z = _59.z;
+    fragColor = texel;
+}
+

gio/shader/gio/zcopy.vert.0.glsl100es

@@ -0,0 +1,35 @@
+#version 100
+
+struct m3x2
+{
+    vec3 r0;
+    vec3 r1;
+};
+
+struct Block
+{
+    vec2 scale;
+    vec2 pos;
+    vec2 uvScale;
+};
+
+uniform Block _block;
+
+varying vec2 vUV;
+attribute vec2 uv;
+attribute vec2 pos;
+
+vec3 transform3x2(m3x2 t, vec3 v)
+{
+    return vec3(dot(t.r0, v), dot(t.r1, v), dot(vec3(0.0, 0.0, 1.0), v));
+}
+
+void main()
+{
+    vUV = vec2(uv * _block.uvScale);
+    vec2 p = vec2((pos * _block.scale) + _block.pos);
+    m3x2 param = m3x2(vec3(1.0, 0.0, 0.0), vec3(0.0, -1.0, 0.0));
+    vec3 param_1 = vec3(p, 0.0);
+    gl_Position = vec4(transform3x2(param, param_1), 1.0);
+}
+

gio/shader/gio/zcopy.vert.0.glsl150

@@ -0,0 +1,35 @@
+#version 150
+
+struct m3x2
+{
+    vec3 r0;
+    vec3 r1;
+};
+
+struct Block
+{
+    vec2 scale;
+    vec2 pos;
+    vec2 uvScale;
+};
+
+uniform Block _block;
+
+out vec2 vUV;
+in vec2 uv;
+in vec2 pos;
+
+vec3 transform3x2(m3x2 t, vec3 v)
+{
+    return vec3(dot(t.r0, v), dot(t.r1, v), dot(vec3(0.0, 0.0, 1.0), v));
+}
+
+void main()
+{
+    vUV = vec2(uv * _block.uvScale);
+    vec2 p = vec2((pos * _block.scale) + _block.pos);
+    m3x2 param = m3x2(vec3(1.0, 0.0, 0.0), vec3(0.0, -1.0, 0.0));
+    vec3 param_1 = vec3(p, 0.0);
+    gl_Position = vec4(transform3x2(param, param_1), 1.0);
+}
+

gio/shader/gio/zcover.frag.0.glsl100es

@@ -0,0 +1,23 @@
+#version 100
+precision mediump float;
+precision highp int;
+
+struct Color
+{
+    vec4 color;
+};
+
+uniform Color _color;
+
+uniform mediump sampler2D cover;
+
+varying highp vec2 vCoverUV;
+varying highp vec2 vUV;
+
+void main()
+{
+    gl_FragData[0] = _color.color;
+    float c = min(abs(texture2D(cover, vCoverUV).x), 1.0);
+    gl_FragData[0] *= c;
+}
+

gio/shader/gio/zcover.frag.0.glsl150

@@ -0,0 +1,22 @@
+#version 150
+
+struct Color
+{
+    vec4 color;
+};
+
+uniform Color _color;
+
+uniform sampler2D cover;
+
+out vec4 fragColor;
+in vec2 vCoverUV;
+in vec2 vUV;
+
+void main()
+{
+    fragColor = _color.color;
+    float c = min(abs(texture(cover, vCoverUV).x), 1.0);
+    fragColor *= c;
+}
+