diff --git a/Cargo.lock b/Cargo.lock
index 1abf289..907a593 100644
--- a/Cargo.lock
+++ b/Cargo.lock
@@ -31,12 +31,6 @@ dependencies = [
  "zerocopy",
 ]
 
-[[package]]
-name = "allocator-api2"
-version = "0.2.18"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "5c6cb57a04249c6480766f7f7cef5467412af1490f8d1e243141daddada3264f"
-
 [[package]]
 name = "android-activity"
 version = "0.6.0"
@@ -44,7 +38,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "ef6978589202a00cd7e118380c448a08b6ed394c3a8df3a430d0898e3a42d046"
 dependencies = [
  "android-properties",
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "cc",
  "cesu8",
  "jni",
@@ -55,7 +49,7 @@ dependencies = [
  "ndk-context",
  "ndk-sys 0.6.0+11769913",
  "num_enum",
- "thiserror",
+ "thiserror 1.0.69",
 ]
 
 [[package]]
@@ -123,18 +117,18 @@ checksum = "ace50bade8e6234aa140d9a2f552bbee1db4d353f69b8217bc503490fc1a9f26"
 
 [[package]]
 name = "bit-set"
-version = "0.6.0"
+version = "0.8.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "f0481a0e032742109b1133a095184ee93d88f3dc9e0d28a5d033dc77a073f44f"
+checksum = "08807e080ed7f9d5433fa9b275196cfc35414f66a0c79d864dc51a0d825231a3"
 dependencies = [
  "bit-vec",
 ]
 
 [[package]]
 name = "bit-vec"
-version = "0.7.0"
+version = "0.8.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "d2c54ff287cfc0a34f38a6b832ea1bd8e448a330b3e40a50859e6488bee07f22"
+checksum = "5e764a1d40d510daf35e07be9eb06e75770908c27d411ee6c92109c9840eaaf7"
 
 [[package]]
 name = "bitflags"
@@ -144,9 +138,12 @@ checksum = "bef38d45163c2f1dde094a7dfd33ccf595c92905c8f8f4fdc18d06fb1037718a"
 
 [[package]]
 name = "bitflags"
-version = "2.6.0"
+version = "2.9.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "b048fb63fd8b5923fc5aa7b340d8e156aec7ec02f0c78fa8a6ddc2613f6f71de"
+checksum = "5c8214115b7bf84099f1309324e63141d4c5d7cc26862f97a0a857dbefe165bd"
+dependencies = [
+ "serde",
+]
 
 [[package]]
 name = "block"
@@ -165,21 +162,35 @@ dependencies = [
 
 [[package]]
 name = "bumpalo"
-version = "3.16.0"
+version = "3.17.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "79296716171880943b8470b5f8d03aa55eb2e645a4874bdbb28adb49162e012c"
+checksum = "1628fb46dfa0b37568d12e5edd512553eccf6a22a78e8bde00bb4aed84d5bdbf"
 
 [[package]]
 name = "bytemuck"
-version = "1.19.0"
+version = "1.22.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "8334215b81e418a0a7bdb8ef0849474f40bb10c8b71f1c4ed315cff49f32494d"
+checksum = "b6b1fc10dbac614ebc03540c9dbd60e83887fda27794998c6528f1782047d540"
+dependencies = [
+ "bytemuck_derive",
+]
+
+[[package]]
+name = "bytemuck_derive"
+version = "1.9.3"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "7ecc273b49b3205b83d648f0690daa588925572cc5063745bfe547fe7ec8e1a1"
+dependencies = [
+ "proc-macro2",
+ "quote",
+ "syn",
+]
 
 [[package]]
 name = "bytes"
-version = "1.8.0"
+version = "1.10.1"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "9ac0150caa2ae65ca5bd83f25c7de183dea78d4d366469f148435e2acfbad0da"
+checksum = "d71b6127be86fdcfddb610f7182ac57211d4b18a3e9c82eb2d17662f2227ad6a"
 
 [[package]]
 name = "calloop"
@@ -187,12 +198,12 @@ version = "0.13.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "b99da2f8558ca23c71f4fd15dc57c906239752dd27ff3c00a1d56b685b7cbfec"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "log",
  "polling",
  "rustix",
  "slab",
- "thiserror",
+ "thiserror 1.0.69",
 ]
 
 [[package]]
@@ -209,9 +220,9 @@ dependencies = [
 
 [[package]]
 name = "cc"
-version = "1.1.31"
+version = "1.2.17"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "c2e7962b54006dcfcc61cb72735f4d89bb97061dd6a7ed882ec6b8ee53714c6f"
+checksum = "1fcb57c740ae1daf453ae85f16e37396f672b039e00d9d866e07ddb24e328e3a"
 dependencies = [
  "jobserver",
  "libc",
@@ -230,12 +241,6 @@ version = "1.0.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "baf1de4339761588bc0619e3cbc0120ee582ebb74b53b4efbf79117bd2da40fd"
 
-[[package]]
-name = "cfg_aliases"
-version = "0.1.1"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "fd16c4719339c4530435d38e511904438d07cce7950afa3718a84ac36c10e89e"
-
 [[package]]
 name = "cfg_aliases"
 version = "0.2.1"
@@ -252,37 +257,6 @@ dependencies = [
  "unicode-width",
 ]
 
-[[package]]
-name = "com"
-version = "0.6.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "7e17887fd17353b65b1b2ef1c526c83e26cd72e74f598a8dc1bee13a48f3d9f6"
-dependencies = [
- "com_macros",
-]
-
-[[package]]
-name = "com_macros"
-version = "0.6.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "d375883580a668c7481ea6631fc1a8863e33cc335bf56bfad8d7e6d4b04b13a5"
-dependencies = [
- "com_macros_support",
- "proc-macro2",
- "syn 1.0.109",
-]
-
-[[package]]
-name = "com_macros_support"
-version = "0.6.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "ad899a1087a9296d5644792d7cb72b8e34c1bec8e7d4fbc002230169a6e8710c"
-dependencies = [
- "proc-macro2",
- "quote",
- "syn 1.0.109",
-]
-
 [[package]]
 name = "combine"
 version = "4.6.7"
@@ -344,9 +318,9 @@ dependencies = [
 
 [[package]]
 name = "crossbeam-utils"
-version = "0.8.20"
+version = "0.8.21"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "22ec99545bb0ed0ea7bb9b8e1e9122ea386ff8a48c0922e43f36d45ab09e0e80"
+checksum = "d0a5c400df2834b80a4c3327b3aad3a4c4cd4de0629063962b03235697506a28"
 
 [[package]]
 name = "cursor-icon"
@@ -354,17 +328,6 @@ version = "1.1.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "96a6ac251f4a2aca6b3f91340350eab87ae57c3f127ffeb585e92bd336717991"
 
-[[package]]
-name = "d3d12"
-version = "22.0.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "bdbd1f579714e3c809ebd822c81ef148b1ceaeb3d535352afc73fd0c4c6a0017"
-dependencies = [
- "bitflags 2.6.0",
- "libloading",
- "winapi",
-]
-
 [[package]]
 name = "dispatch"
 version = "0.2.0"
@@ -382,9 +345,9 @@ dependencies = [
 
 [[package]]
 name = "document-features"
-version = "0.2.10"
+version = "0.2.11"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "cb6969eaabd2421f8a2775cfd2471a2b634372b4a25d41e3bd647b79912850a0"
+checksum = "95249b50c6c185bee49034bcb378a49dc2b5dff0be90ff6616d31d64febab05d"
 dependencies = [
  "litrs",
 ]
@@ -403,20 +366,26 @@ checksum = "f25c0e292a7ca6d6498557ff1df68f32c99850012b6ea401cf8daf771f22ff53"
 
 [[package]]
 name = "equivalent"
-version = "1.0.1"
+version = "1.0.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "5443807d6dff69373d433ab9ef5378ad8df50ca6298caf15de6e52e24aaf54d5"
+checksum = "877a4ace8713b0bcf2a4e7eec82529c029f1d0619886d18145fea96c3ffe5c0f"
 
 [[package]]
 name = "errno"
-version = "0.3.9"
+version = "0.3.10"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "534c5cf6194dfab3db3242765c03bbe257cf92f22b38f6bc0c58d59108a820ba"
+checksum = "33d852cb9b869c2a9b3df2f71a3074817f01e1844f839a144f5fcef059a4eb5d"
 dependencies = [
  "libc",
- "windows-sys 0.52.0",
+ "windows-sys 0.59.0",
 ]
 
+[[package]]
+name = "foldhash"
+version = "0.1.5"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "d9c4f5dac5e15c24eb999c26181a6ca40b39fe946cbe4c263c7209467bc83af2"
+
 [[package]]
 name = "foreign-types"
 version = "0.5.0"
@@ -435,7 +404,7 @@ checksum = "1a5c6c585bc94aaf2c7b51dd4c2ba22680844aba4c687be581871a6f518c5742"
 dependencies = [
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
 ]
 
 [[package]]
@@ -490,9 +459,9 @@ dependencies = [
 
 [[package]]
 name = "glow"
-version = "0.13.1"
+version = "0.16.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "bd348e04c43b32574f2de31c8bb397d96c9fcfa1371bd4ca6d8bdc464ab121b1"
+checksum = "c5e5ea60d70410161c8bf5da3fdfeaa1c72ed2c15f8bbb9d19fe3a4fad085f08"
 dependencies = [
  "js-sys",
  "slotmap",
@@ -502,9 +471,9 @@ dependencies = [
 
 [[package]]
 name = "glutin_wgl_sys"
-version = "0.6.0"
+version = "0.6.1"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "0a4e1951bbd9434a81aa496fe59ccc2235af3820d27b85f9314e279609211e2c"
+checksum = "2c4ee00b289aba7a9e5306d57c2d05499b2e5dc427f84ac708bd2c090212cf3e"
 dependencies = [
  "gl_generator",
 ]
@@ -515,7 +484,7 @@ version = "0.6.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "fbcd2dba93594b227a1f57ee09b8b9da8892c34d55aa332e034a228d0fe6a171"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "gpu-alloc-types",
 ]
 
@@ -525,31 +494,30 @@ version = "0.3.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "98ff03b468aa837d70984d55f5d3f846f6ec31fe34bbb97c4f85219caeee1ca4"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
 ]
 
 [[package]]
 name = "gpu-allocator"
-version = "0.26.0"
+version = "0.27.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "fdd4240fc91d3433d5e5b0fc5b67672d771850dc19bbee03c1381e19322803d7"
+checksum = "c151a2a5ef800297b4e79efa4f4bec035c5f51d5ae587287c9b952bdf734cacd"
 dependencies = [
  "log",
  "presser",
- "thiserror",
- "winapi",
+ "thiserror 1.0.69",
  "windows",
 ]
 
 [[package]]
 name = "gpu-descriptor"
-version = "0.3.0"
+version = "0.3.1"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "9c08c1f623a8d0b722b8b99f821eb0ba672a1618f0d3b16ddbee1cedd2dd8557"
+checksum = "dcf29e94d6d243368b7a56caa16bc213e4f9f8ed38c4d9557069527b5d5281ca"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "gpu-descriptor-types",
- "hashbrown 0.14.5",
+ "hashbrown",
 ]
 
 [[package]]
@@ -558,39 +526,23 @@ version = "0.2.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "fdf242682df893b86f33a73828fb09ca4b2d3bb6cc95249707fc684d27484b91"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
 ]
 
 [[package]]
 name = "hashbrown"
-version = "0.14.5"
+version = "0.15.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "e5274423e17b7c9fc20b6e7e208532f9b19825d82dfd615708b70edd83df41f1"
+checksum = "bf151400ff0baff5465007dd2f3e717f3fe502074ca563069ce3a6629d07b289"
 dependencies = [
- "ahash",
- "allocator-api2",
+ "foldhash",
 ]
 
 [[package]]
-name = "hashbrown"
-version = "0.15.0"
+name = "heck"
+version = "0.5.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "1e087f84d4f86bf4b218b927129862374b72199ae7d8657835f1e89000eea4fb"
-
-[[package]]
-name = "hassle-rs"
-version = "0.11.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "af2a7e73e1f34c48da31fb668a907f250794837e08faa144fd24f0b8b741e890"
-dependencies = [
- "bitflags 2.6.0",
- "com",
- "libc",
- "libloading",
- "thiserror",
- "widestring",
- "winapi",
-]
+checksum = "2304e00983f87ffb38b55b444b5e3b60a884b5d30c0fca7d82fe33449bbe55ea"
 
 [[package]]
 name = "hermit-abi"
@@ -606,12 +558,12 @@ checksum = "dfa686283ad6dd069f105e5ab091b04c62850d3e4cf5d67debad1933f55023df"
 
 [[package]]
 name = "indexmap"
-version = "2.6.0"
+version = "2.8.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "707907fe3c25f5424cce2cb7e1cbcafee6bdbe735ca90ef77c29e84591e5b9da"
+checksum = "3954d50fe15b02142bf25d3b8bdadb634ec3948f103d04ffe3031bc8fe9d7058"
 dependencies = [
  "equivalent",
- "hashbrown 0.15.0",
+ "hashbrown",
 ]
 
 [[package]]
@@ -625,7 +577,7 @@ dependencies = [
  "combine",
  "jni-sys",
  "log",
- "thiserror",
+ "thiserror 1.0.69",
  "walkdir",
  "windows-sys 0.45.0",
 ]
@@ -647,10 +599,11 @@ dependencies = [
 
 [[package]]
 name = "js-sys"
-version = "0.3.72"
+version = "0.3.77"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "6a88f1bda2bd75b0452a14784937d796722fdebfe50df998aeb3f0b7603019a9"
+checksum = "1cfaf33c695fc6e08064efbc1f72ec937429614f25eef83af942d0e227c3a28f"
 dependencies = [
+ "once_cell",
  "wasm-bindgen",
 ]
 
@@ -673,15 +626,15 @@ checksum = "e2db585e1d738fc771bf08a151420d3ed193d9d895a36df7f6f8a9456b911ddc"
 
 [[package]]
 name = "libc"
-version = "0.2.161"
+version = "0.2.171"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "8e9489c2807c139ffd9c1794f4af0ebe86a828db53ecdc7fea2111d0fed085d1"
+checksum = "c19937216e9d3aa9956d9bb8dfc0b0c8beb6058fc4f7a4dc4d850edf86a237d6"
 
 [[package]]
 name = "libloading"
-version = "0.8.5"
+version = "0.8.6"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "4979f22fdb869068da03c9f7528f8297c6fd2606bc3a4affe42e6a823fdb8da4"
+checksum = "fc2f4eb4bc735547cfed7c0a4922cbd04a4655978c09b54f1f7b228750664c34"
 dependencies = [
  "cfg-if",
  "windows-targets 0.52.6",
@@ -693,16 +646,16 @@ version = "0.1.3"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "c0ff37bd590ca25063e35af745c343cb7a0271906fb7b37e4813e8f79f00268d"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "libc",
- "redox_syscall 0.5.7",
+ "redox_syscall 0.5.10",
 ]
 
 [[package]]
 name = "linux-raw-sys"
-version = "0.4.14"
+version = "0.4.15"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "78b3ae25bc7c8c38cec158d1f2757ee79e9b3740fbc7ccf0e59e4b08d793fa89"
+checksum = "d26c52dbd32dccf2d10cac7725f8eae5296885fb5703b261f7d0a0739ec807ab"
 
 [[package]]
 name = "litrs"
@@ -722,9 +675,9 @@ dependencies = [
 
 [[package]]
 name = "log"
-version = "0.4.22"
+version = "0.4.27"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "a7a70ba024b9dc04c27ea2f0c0548feb474ec5c54bba33a7f72f873a39d07b24"
+checksum = "13dc2df351e3202783a1fe0d44375f7295ffb4049267b0f3018346dc122a1d94"
 
 [[package]]
 name = "malloc_buf"
@@ -762,11 +715,11 @@ dependencies = [
 
 [[package]]
 name = "metal"
-version = "0.29.0"
+version = "0.31.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "7ecfd3296f8c56b7c1f6fbac3c71cefa9d78ce009850c45000015f206dc7fa21"
+checksum = "f569fb946490b5743ad69813cb19629130ce9374034abe31614a36402d18f99e"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block",
  "core-graphics-types",
  "foreign-types",
@@ -777,30 +730,31 @@ dependencies = [
 
 [[package]]
 name = "naga"
-version = "22.1.0"
+version = "24.0.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "8bd5a652b6faf21496f2cfd88fc49989c8db0825d1f6746b1a71a6ede24a63ad"
+checksum = "e380993072e52eef724eddfcde0ed013b0c023c3f0417336ed041aa9f076994e"
 dependencies = [
  "arrayvec",
  "bit-set",
- "bitflags 2.6.0",
- "cfg_aliases 0.1.1",
+ "bitflags 2.9.0",
+ "cfg_aliases",
  "codespan-reporting",
  "hexf-parse",
  "indexmap",
  "log",
  "rustc-hash",
  "spirv",
+ "strum",
  "termcolor",
- "thiserror",
+ "thiserror 2.0.12",
  "unicode-xid",
 ]
 
 [[package]]
 name = "nalgebra"
-version = "0.33.1"
+version = "0.33.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "3bf139e93ad757869338ad85239cb1d6c067b23b94e5846e637ca6328ee4be60"
+checksum = "26aecdf64b707efd1310e3544d709c5c0ac61c13756046aaaba41be5c4f66a3b"
 dependencies = [
  "approx",
  "matrixmultiply",
@@ -820,7 +774,7 @@ checksum = "254a5372af8fc138e36684761d3c0cdb758a4410e938babcff1c860ce14ddbfc"
 dependencies = [
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
 ]
 
 [[package]]
@@ -829,13 +783,13 @@ version = "0.9.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "c3f42e7bbe13d351b6bead8286a43aac9534b82bd3cc43e47037f012ebfd62d4"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "jni-sys",
  "log",
  "ndk-sys 0.6.0+11769913",
  "num_enum",
  "raw-window-handle",
- "thiserror",
+ "thiserror 1.0.69",
 ]
 
 [[package]]
@@ -928,7 +882,7 @@ dependencies = [
  "proc-macro-crate",
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
 ]
 
 [[package]]
@@ -962,7 +916,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "e4e89ad9e3d7d297152b17d39ed92cd50ca8063a89a9fa569046d41568891eff"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "libc",
  "objc2",
@@ -978,7 +932,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "74dd3b56391c7a0596a295029734d3c1c5e7e510a4cb30245f8221ccea96b009"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "objc2",
  "objc2-core-location",
@@ -1002,7 +956,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "617fbf49e071c178c0b24c080767db52958f716d9eabdf0890523aeae54773ef"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "objc2",
  "objc2-foundation",
@@ -1034,9 +988,9 @@ dependencies = [
 
 [[package]]
 name = "objc2-encode"
-version = "4.0.3"
+version = "4.1.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "7891e71393cd1f227313c9379a26a584ff3d7e6e7159e988851f0934c993f0f8"
+checksum = "ef25abbcd74fb2609453eb695bd2f860d389e457f67dc17cafc8b8cbc89d0c33"
 
 [[package]]
 name = "objc2-foundation"
@@ -1044,7 +998,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "0ee638a5da3799329310ad4cfa62fbf045d5f56e3ef5ba4149e7452dcf89d5a8"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "dispatch",
  "libc",
@@ -1069,7 +1023,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "dd0cba1276f6023976a406a14ffa85e1fdd19df6b0f737b063b95f6c8c7aadd6"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "objc2",
  "objc2-foundation",
@@ -1081,7 +1035,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "e42bee7bff906b14b167da2bac5efe6b6a07e6f7c0a21a7308d40c960242dc7a"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "objc2",
  "objc2-foundation",
@@ -1104,7 +1058,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "b8bb46798b20cd6b91cbd113524c490f1686f4c4e8f49502431415f3512e2b6f"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "objc2",
  "objc2-cloud-kit",
@@ -1136,7 +1090,7 @@ version = "0.2.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "76cfcbf642358e8689af64cee815d139339f3ed8ad05103ed5eaf73db8d84cb3"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "objc2",
  "objc2-core-location",
@@ -1145,9 +1099,9 @@ dependencies = [
 
 [[package]]
 name = "once_cell"
-version = "1.20.2"
+version = "1.21.3"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "1261fe7e33c73b354eab43b1273a57c8f967d0391e80353e51f764ac02cf6775"
+checksum = "42f5e15c9953c5e4ccceeb2e7382a716482c34515315f7b03532b8b4e8393d2d"
 
 [[package]]
 name = "orbclient"
@@ -1158,6 +1112,15 @@ dependencies = [
  "libredox",
 ]
 
+[[package]]
+name = "ordered-float"
+version = "4.6.0"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "7bb71e1b3fa6ca1c61f383464aaf2bb0e2f8e772a1f01d486832464de363b951"
+dependencies = [
+ "num-traits",
+]
+
 [[package]]
 name = "owned_ttf_parser"
 version = "0.25.0"
@@ -1185,7 +1148,7 @@ checksum = "1e401f977ab385c9e4e3ab30627d6f26d00e2c73eef317493c4ec6d468726cf8"
 dependencies = [
  "cfg-if",
  "libc",
- "redox_syscall 0.5.7",
+ "redox_syscall 0.5.10",
  "smallvec",
  "windows-targets 0.52.6",
 ]
@@ -1204,41 +1167,41 @@ checksum = "e3148f5046208a5d56bcfc03053e3ca6334e51da8dfb19b6cdc8b306fae3283e"
 
 [[package]]
 name = "pin-project"
-version = "1.1.6"
+version = "1.1.10"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "baf123a161dde1e524adf36f90bc5d8d3462824a9c43553ad07a8183161189ec"
+checksum = "677f1add503faace112b9f1373e43e9e054bfdd22ff1a63c1bc485eaec6a6a8a"
 dependencies = [
  "pin-project-internal",
 ]
 
 [[package]]
 name = "pin-project-internal"
-version = "1.1.6"
+version = "1.1.10"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "a4502d8515ca9f32f1fb543d987f63d95a14934883db45bdb48060b6b69257f8"
+checksum = "6e918e4ff8c4549eb882f14b3a4bc8c8bc93de829416eacf579f1207a8fbf861"
 dependencies = [
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
 ]
 
 [[package]]
 name = "pin-project-lite"
-version = "0.2.14"
+version = "0.2.16"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "bda66fc9667c18cb2758a2ac84d1167245054bcf85d5d1aaa6923f45801bdd02"
+checksum = "3b3cff922bd51709b605d9ead9aa71031d81447142d828eb4a6eba76fe619f9b"
 
 [[package]]
 name = "pkg-config"
-version = "0.3.31"
+version = "0.3.32"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "953ec861398dccce10c670dfeaf3ec4911ca479e9c02154b3a215178c5f566f2"
+checksum = "7edddbd0b52d732b21ad9a5fab5c704c14cd949e5e9a1ec5929a24fded1b904c"
 
 [[package]]
 name = "polling"
-version = "3.7.3"
+version = "3.7.4"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "cc2790cd301dec6cd3b7a025e4815cf825724a51c98dccfe6a3e55f05ffb6511"
+checksum = "a604568c3202727d1507653cb121dbd627a58684eb09a820fd746bee38b4442f"
 dependencies = [
  "cfg-if",
  "concurrent-queue",
@@ -1263,18 +1226,18 @@ checksum = "e8cf8e6a8aa66ce33f63993ffc4ea4271eb5b0530a9002db8455ea6050c77bfa"
 
 [[package]]
 name = "proc-macro-crate"
-version = "3.2.0"
+version = "3.3.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "8ecf48c7ca261d60b74ab1a7b20da18bede46776b2e55535cb958eb595c5fa7b"
+checksum = "edce586971a4dfaa28950c6f18ed55e0406c1ab88bbce2c6f6293a7aaba73d35"
 dependencies = [
  "toml_edit",
 ]
 
 [[package]]
 name = "proc-macro2"
-version = "1.0.89"
+version = "1.0.94"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "f139b0662de085916d1fb67d2b4169d1addddda1919e696f3252b740b629986e"
+checksum = "a31971752e70b8b2686d7e46ec17fb38dad4051d94024c88df49b667caea9c84"
 dependencies = [
  "unicode-ident",
 ]
@@ -1287,27 +1250,27 @@ checksum = "afbdc74edc00b6f6a218ca6a5364d6226a259d4b8ea1af4a0ea063f27e179f4d"
 
 [[package]]
 name = "quick-xml"
-version = "0.36.2"
+version = "0.37.3"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "f7649a7b4df05aed9ea7ec6f628c67c9953a43869b8bc50929569b2999d443fe"
+checksum = "bf763ab1c7a3aa408be466efc86efe35ed1bd3dd74173ed39d6b0d0a6f0ba148"
 dependencies = [
  "memchr",
 ]
 
 [[package]]
 name = "quote"
-version = "1.0.37"
+version = "1.0.40"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "b5b9d34b8991d19d98081b46eacdd8eb58c6f2b201139f7c5f643cc155a633af"
+checksum = "1885c039570dc00dcb4ff087a89e185fd56bae234ddc7f056a945bf36467248d"
 dependencies = [
  "proc-macro2",
 ]
 
 [[package]]
 name = "range-alloc"
-version = "0.1.3"
+version = "0.1.4"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "9c8a99fddc9f0ba0a85884b8d14e3592853e787d581ca1816c91349b10e4eeab"
+checksum = "c3d6831663a5098ea164f89cff59c6284e95f4e3c76ce9848d4529f5ccca9bde"
 
 [[package]]
 name = "raw-window-handle"
@@ -1332,11 +1295,11 @@ dependencies = [
 
 [[package]]
 name = "redox_syscall"
-version = "0.5.7"
+version = "0.5.10"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "9b6dfecf2c74bce2466cabf93f6664d6998a69eb21e39f4207930065b27b771f"
+checksum = "0b8c0c260b63a8219631167be35e6a988e9554dbd323f8bd08439c8ed1302bd1"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
 ]
 
 [[package]]
@@ -1353,22 +1316,28 @@ checksum = "08d43f7aa6b08d49f382cde6a7982047c3426db949b1424bc4b7ec9ae12c6ce2"
 
 [[package]]
 name = "rustix"
-version = "0.38.37"
+version = "0.38.44"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "8acb788b847c24f28525660c4d7758620a7210875711f79e7f663cc152726811"
+checksum = "fdb5bc1ae2baa591800df16c9ca78619bf65c0488b41b96ccec5d11220d8c154"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "errno",
  "libc",
  "linux-raw-sys",
- "windows-sys 0.52.0",
+ "windows-sys 0.59.0",
 ]
 
 [[package]]
-name = "safe_arch"
-version = "0.7.2"
+name = "rustversion"
+version = "1.0.20"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "c3460605018fdc9612bce72735cba0d27efbcd9904780d44c7e3a9948f96148a"
+checksum = "eded382c5f5f786b989652c49544c4877d9f015cc22e145a5ea8ea66c2921cd2"
+
+[[package]]
+name = "safe_arch"
+version = "0.7.4"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "96b02de82ddbe1b636e6170c21be622223aea188ef2e139be0a5b219ec215323"
 dependencies = [
  "bytemuck",
 ]
@@ -1409,22 +1378,22 @@ dependencies = [
 
 [[package]]
 name = "serde"
-version = "1.0.213"
+version = "1.0.219"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "3ea7893ff5e2466df8d720bb615088341b295f849602c6956047f8f80f0e9bc1"
+checksum = "5f0e2c6ed6606019b4e29e69dbaba95b11854410e5347d525002456dbbb786b6"
 dependencies = [
  "serde_derive",
 ]
 
 [[package]]
 name = "serde_derive"
-version = "1.0.213"
+version = "1.0.219"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "7e85ad2009c50b58e87caa8cd6dac16bdf511bbfb7af6c33df902396aa480fa5"
+checksum = "5b0276cf7f2c73365f7157c8123c21cd9a50fbbd844757af28ca1f5925fc2a00"
 dependencies = [
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
 ]
 
 [[package]]
@@ -1466,9 +1435,9 @@ dependencies = [
 
 [[package]]
 name = "smallvec"
-version = "1.13.2"
+version = "1.14.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "3c5e1a9a646d36c3599cd173a41282daf47c44583ad367b8e6837255952e5c67"
+checksum = "7fcf8323ef1faaee30a44a340193b1ac6814fd9b7b4e88e9d4519a3e4abe1cfd"
 
 [[package]]
 name = "smithay-client-toolkit"
@@ -1476,7 +1445,7 @@ version = "0.19.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "3457dea1f0eb631b4034d61d4d8c32074caa6cd1ab2d59f2327bd8461e2c0016"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "calloop",
  "calloop-wayland-source",
  "cursor-icon",
@@ -1484,7 +1453,7 @@ dependencies = [
  "log",
  "memmap2",
  "rustix",
- "thiserror",
+ "thiserror 1.0.69",
  "wayland-backend",
  "wayland-client",
  "wayland-csd-frame",
@@ -1510,7 +1479,7 @@ version = "0.3.0+sdk-1.3.268.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "eda41003dc44290527a59b13432d4a0379379fa074b70174882adfbdfd917844"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
 ]
 
 [[package]]
@@ -1526,21 +1495,32 @@ source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "6637bab7722d379c8b41ba849228d680cc12d0a45ba1fa2b48f2a30577a06731"
 
 [[package]]
-name = "syn"
-version = "1.0.109"
+name = "strum"
+version = "0.26.3"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "72b64191b275b66ffe2469e8af2c1cfe3bafa67b529ead792a6d0160888b4237"
+checksum = "8fec0f0aef304996cf250b31b5a10dee7980c85da9d759361292b8bca5a18f06"
 dependencies = [
+ "strum_macros",
+]
+
+[[package]]
+name = "strum_macros"
+version = "0.26.4"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "4c6bee85a5a24955dc440386795aa378cd9cf82acd5f764469152d2270e581be"
+dependencies = [
+ "heck",
  "proc-macro2",
  "quote",
- "unicode-ident",
+ "rustversion",
+ "syn",
 ]
 
 [[package]]
 name = "syn"
-version = "2.0.84"
+version = "2.0.100"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "6a2c4efbc0b0670e3d41f388e3cb936ff364bf681703b4c92ae26ca509966111"
+checksum = "b09a44accad81e1ba1cd74a32461ba89dee89095ba17b32f5d03683b1b1fc2a0"
 dependencies = [
  "proc-macro2",
  "quote",
@@ -1558,22 +1538,42 @@ dependencies = [
 
 [[package]]
 name = "thiserror"
-version = "1.0.65"
+version = "1.0.69"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "5d11abd9594d9b38965ef50805c5e469ca9cc6f197f883f717e0269a3057b3d5"
+checksum = "b6aaf5339b578ea85b50e080feb250a3e8ae8cfcdff9a461c9ec2904bc923f52"
 dependencies = [
- "thiserror-impl",
+ "thiserror-impl 1.0.69",
+]
+
+[[package]]
+name = "thiserror"
+version = "2.0.12"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "567b8a2dae586314f7be2a752ec7474332959c6460e02bde30d702a66d488708"
+dependencies = [
+ "thiserror-impl 2.0.12",
 ]
 
 [[package]]
 name = "thiserror-impl"
-version = "1.0.65"
+version = "1.0.69"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "ae71770322cbd277e69d762a16c444af02aa0575ac0d174f0b9562d3b37f8602"
+checksum = "4fee6c4efc90059e10f81e6d42c60a18f76588c3d74cb83a0b242a2b6c7504c1"
 dependencies = [
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
+]
+
+[[package]]
+name = "thiserror-impl"
+version = "2.0.12"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "7f7cf42b4507d8ea322120659672cf1b9dbb93f8f2d4ecfd6e51350ff5b17a1d"
+dependencies = [
+ "proc-macro2",
+ "quote",
+ "syn",
 ]
 
 [[package]]
@@ -1609,9 +1609,9 @@ checksum = "0dd7358ecb8fc2f8d014bf86f6f638ce72ba252a2c3a2572f2a795f1d23efb41"
 
 [[package]]
 name = "toml_edit"
-version = "0.22.22"
+version = "0.22.24"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "4ae48d6208a266e853d946088ed816055e556cc6028c5e8e2b84d9fa5dd7c7f5"
+checksum = "17b4795ff5edd201c7cd6dca065ae59972ce77d1b80fa0a84d94950ece7d1474"
 dependencies = [
  "indexmap",
  "toml_datetime",
@@ -1620,9 +1620,9 @@ dependencies = [
 
 [[package]]
 name = "tracing"
-version = "0.1.40"
+version = "0.1.41"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "c3523ab5a71916ccf420eebdf5521fcef02141234bbc0b8a49f2fdc4544364ef"
+checksum = "784e0ac535deb450455cbfa28a6f0df145ea1bb7ae51b821cf5e7927fdcfbdd0"
 dependencies = [
  "pin-project-lite",
  "tracing-core",
@@ -1630,27 +1630,27 @@ dependencies = [
 
 [[package]]
 name = "tracing-core"
-version = "0.1.32"
+version = "0.1.33"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "c06d3da6113f116aaee68e4d601191614c9053067f9ab7f6edbcb161237daa54"
+checksum = "e672c95779cf947c5311f83787af4fa8fffd12fb27e4993211a84bdfd9610f9c"
 
 [[package]]
 name = "ttf-parser"
-version = "0.25.0"
+version = "0.25.1"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "5902c5d130972a0000f60860bfbf46f7ca3db5391eddfedd1b8728bd9dc96c0e"
+checksum = "d2df906b07856748fa3f6e0ad0cbaa047052d4a7dd609e231c4f72cee8c36f31"
 
 [[package]]
 name = "typenum"
-version = "1.17.0"
+version = "1.18.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "42ff0bf0c66b8238c6f3b578df37d0b7848e55df8577b3f74f92a69acceeb825"
+checksum = "1dccffe3ce07af9386bfd29e80c0ab1a8205a2fc34e4bcd40364df902cfa8f3f"
 
 [[package]]
 name = "unicode-ident"
-version = "1.0.13"
+version = "1.0.18"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "e91b56cd4cadaeb79bbf1a5645f6b4f8dc5bde8834ad5894a8db35fda9efa1fe"
+checksum = "5a5f39404a5da50712a4c1eecf25e90dd62b613502b7e925fd4e4d19b5c96512"
 
 [[package]]
 name = "unicode-segmentation"
@@ -1694,47 +1694,48 @@ checksum = "9c8d87e72b64a3b4db28d11ce29237c246188f4f51057d65a7eab63b7987e423"
 
 [[package]]
 name = "wasm-bindgen"
-version = "0.2.95"
+version = "0.2.100"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "128d1e363af62632b8eb57219c8fd7877144af57558fb2ef0368d0087bddeb2e"
+checksum = "1edc8929d7499fc4e8f0be2262a241556cfc54a0bea223790e71446f2aab1ef5"
 dependencies = [
  "cfg-if",
  "once_cell",
+ "rustversion",
  "wasm-bindgen-macro",
 ]
 
 [[package]]
 name = "wasm-bindgen-backend"
-version = "0.2.95"
+version = "0.2.100"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "cb6dd4d3ca0ddffd1dd1c9c04f94b868c37ff5fac97c30b97cff2d74fce3a358"
+checksum = "2f0a0651a5c2bc21487bde11ee802ccaf4c51935d0d3d42a6101f98161700bc6"
 dependencies = [
  "bumpalo",
  "log",
- "once_cell",
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
  "wasm-bindgen-shared",
 ]
 
 [[package]]
 name = "wasm-bindgen-futures"
-version = "0.4.45"
+version = "0.4.50"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "cc7ec4f8827a71586374db3e87abdb5a2bb3a15afed140221307c3ec06b1f63b"
+checksum = "555d470ec0bc3bb57890405e5d4322cc9ea83cebb085523ced7be4144dac1e61"
 dependencies = [
  "cfg-if",
  "js-sys",
+ "once_cell",
  "wasm-bindgen",
  "web-sys",
 ]
 
 [[package]]
 name = "wasm-bindgen-macro"
-version = "0.2.95"
+version = "0.2.100"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "e79384be7f8f5a9dd5d7167216f022090cf1f9ec128e6e6a482a2cb5c5422c56"
+checksum = "7fe63fc6d09ed3792bd0897b314f53de8e16568c2b3f7982f468c0bf9bd0b407"
 dependencies = [
  "quote",
  "wasm-bindgen-macro-support",
@@ -1742,28 +1743,31 @@ dependencies = [
 
 [[package]]
 name = "wasm-bindgen-macro-support"
-version = "0.2.95"
+version = "0.2.100"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "26c6ab57572f7a24a4985830b120de1594465e5d500f24afe89e16b4e833ef68"
+checksum = "8ae87ea40c9f689fc23f209965b6fb8a99ad69aeeb0231408be24920604395de"
 dependencies = [
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
  "wasm-bindgen-backend",
  "wasm-bindgen-shared",
 ]
 
 [[package]]
 name = "wasm-bindgen-shared"
-version = "0.2.95"
+version = "0.2.100"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "65fc09f10666a9f147042251e0dda9c18f166ff7de300607007e96bdebc1068d"
+checksum = "1a05d73b933a847d6cccdda8f838a22ff101ad9bf93e33684f39c1f5f0eece3d"
+dependencies = [
+ "unicode-ident",
+]
 
 [[package]]
 name = "wayland-backend"
-version = "0.3.7"
+version = "0.3.8"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "056535ced7a150d45159d3a8dc30f91a2e2d588ca0b23f70e56033622b8016f6"
+checksum = "b7208998eaa3870dad37ec8836979581506e0c5c64c20c9e79e9d2a10d6f47bf"
 dependencies = [
  "cc",
  "downcast-rs",
@@ -1775,11 +1779,11 @@ dependencies = [
 
 [[package]]
 name = "wayland-client"
-version = "0.31.7"
+version = "0.31.8"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "b66249d3fc69f76fd74c82cc319300faa554e9d865dab1f7cd66cc20db10b280"
+checksum = "c2120de3d33638aaef5b9f4472bff75f07c56379cf76ea320bd3a3d65ecaf73f"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "rustix",
  "wayland-backend",
  "wayland-scanner",
@@ -1791,16 +1795,16 @@ version = "0.3.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "625c5029dbd43d25e6aa9615e88b829a5cad13b2819c4ae129fdbb7c31ab4c7e"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "cursor-icon",
  "wayland-backend",
 ]
 
 [[package]]
 name = "wayland-cursor"
-version = "0.31.7"
+version = "0.31.8"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "32b08bc3aafdb0035e7fe0fdf17ba0c09c268732707dca4ae098f60cb28c9e4c"
+checksum = "a93029cbb6650748881a00e4922b076092a6a08c11e7fbdb923f064b23968c5d"
 dependencies = [
  "rustix",
  "wayland-client",
@@ -1809,11 +1813,11 @@ dependencies = [
 
 [[package]]
 name = "wayland-protocols"
-version = "0.32.5"
+version = "0.32.6"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "7cd0ade57c4e6e9a8952741325c30bf82f4246885dca8bf561898b86d0c1f58e"
+checksum = "0781cf46869b37e36928f7b432273c0995aa8aed9552c556fb18754420541efc"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "wayland-backend",
  "wayland-client",
  "wayland-scanner",
@@ -1821,11 +1825,11 @@ dependencies = [
 
 [[package]]
 name = "wayland-protocols-plasma"
-version = "0.3.5"
+version = "0.3.6"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "9b31cab548ee68c7eb155517f2212049dc151f7cd7910c2b66abfd31c3ee12bd"
+checksum = "7ccaacc76703fefd6763022ac565b590fcade92202492381c95b2edfdf7d46b3"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "wayland-backend",
  "wayland-client",
  "wayland-protocols",
@@ -1834,11 +1838,11 @@ dependencies = [
 
 [[package]]
 name = "wayland-protocols-wlr"
-version = "0.3.5"
+version = "0.3.6"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "782e12f6cd923c3c316130d56205ebab53f55d6666b7faddfad36cecaeeb4022"
+checksum = "248a02e6f595aad796561fa82d25601bd2c8c3b145b1c7453fc8f94c1a58f8b2"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "wayland-backend",
  "wayland-client",
  "wayland-protocols",
@@ -1847,9 +1851,9 @@ dependencies = [
 
 [[package]]
 name = "wayland-scanner"
-version = "0.31.5"
+version = "0.31.6"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "597f2001b2e5fc1121e3d5b9791d3e78f05ba6bfa4641053846248e3a13661c3"
+checksum = "896fdafd5d28145fce7958917d69f2fd44469b1d4e861cb5961bcbeebc6d1484"
 dependencies = [
  "proc-macro2",
  "quick-xml",
@@ -1858,9 +1862,9 @@ dependencies = [
 
 [[package]]
 name = "wayland-sys"
-version = "0.31.5"
+version = "0.31.6"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "efa8ac0d8e8ed3e3b5c9fc92c7881406a268e11555abe36493efabe649a29e09"
+checksum = "dbcebb399c77d5aa9fa5db874806ee7b4eba4e73650948e8f93963f128896615"
 dependencies = [
  "dlib",
  "log",
@@ -1870,9 +1874,9 @@ dependencies = [
 
 [[package]]
 name = "web-sys"
-version = "0.3.72"
+version = "0.3.77"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "f6488b90108c040df0fe62fa815cbdee25124641df01814dd7282749234c6112"
+checksum = "33b6dd2ef9186f1f2072e409e99cd22a975331a6b3591b12c764e0e55c60d5d2"
 dependencies = [
  "js-sys",
  "wasm-bindgen",
@@ -1890,12 +1894,13 @@ dependencies = [
 
 [[package]]
 name = "wgpu"
-version = "22.1.0"
+version = "24.0.3"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "e1d1c4ba43f80542cf63a0a6ed3134629ae73e8ab51e4b765a67f3aa062eb433"
+checksum = "35904fb00ba2d2e0a4d002fcbbb6e1b89b574d272a50e5fc95f6e81cf281c245"
 dependencies = [
  "arrayvec",
- "cfg_aliases 0.1.1",
+ "bitflags 2.9.0",
+ "cfg_aliases",
  "document-features",
  "js-sys",
  "log",
@@ -1915,14 +1920,14 @@ dependencies = [
 
 [[package]]
 name = "wgpu-core"
-version = "22.1.0"
+version = "24.0.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "0348c840d1051b8e86c3bcd31206080c5e71e5933dabd79be1ce732b0b2f089a"
+checksum = "671c25545d479b47d3f0a8e373aceb2060b67c6eb841b24ac8c32348151c7a0c"
 dependencies = [
  "arrayvec",
  "bit-vec",
- "bitflags 2.6.0",
- "cfg_aliases 0.1.1",
+ "bitflags 2.9.0",
+ "cfg_aliases",
  "document-features",
  "indexmap",
  "log",
@@ -1933,32 +1938,31 @@ dependencies = [
  "raw-window-handle",
  "rustc-hash",
  "smallvec",
- "thiserror",
+ "thiserror 2.0.12",
  "wgpu-hal",
  "wgpu-types",
 ]
 
 [[package]]
 name = "wgpu-hal"
-version = "22.0.0"
+version = "24.0.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "f6bbf4b4de8b2a83c0401d9e5ae0080a2792055f25859a02bf9be97952bbed4f"
+checksum = "4317a17171dc20e6577bf606796794580accae0716a69edbc7388c86a3ec9f23"
 dependencies = [
  "android_system_properties",
  "arrayvec",
  "ash",
  "bit-set",
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block",
- "cfg_aliases 0.1.1",
+ "bytemuck",
+ "cfg_aliases",
  "core-graphics-types",
- "d3d12",
  "glow",
  "glutin_wgl_sys",
  "gpu-alloc",
  "gpu-allocator",
  "gpu-descriptor",
- "hassle-rs",
  "js-sys",
  "khronos-egl",
  "libc",
@@ -1969,6 +1973,7 @@ dependencies = [
  "ndk-sys 0.5.0+25.2.9519653",
  "objc",
  "once_cell",
+ "ordered-float",
  "parking_lot",
  "profiling",
  "range-alloc",
@@ -1976,56 +1981,36 @@ dependencies = [
  "renderdoc-sys",
  "rustc-hash",
  "smallvec",
- "thiserror",
+ "thiserror 2.0.12",
  "wasm-bindgen",
  "web-sys",
  "wgpu-types",
- "winapi",
+ "windows",
+ "windows-core",
 ]
 
 [[package]]
 name = "wgpu-types"
-version = "22.0.0"
+version = "24.0.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "bc9d91f0e2c4b51434dfa6db77846f2793149d8e73f800fa2e41f52b8eac3c5d"
+checksum = "50ac044c0e76c03a0378e7786ac505d010a873665e2d51383dcff8dd227dc69c"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "js-sys",
+ "log",
  "web-sys",
 ]
 
 [[package]]
 name = "wide"
-version = "0.7.28"
+version = "0.7.32"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "b828f995bf1e9622031f8009f8481a85406ce1f4d4588ff746d872043e855690"
+checksum = "41b5576b9a81633f3e8df296ce0063042a73507636cbe956c61133dd7034ab22"
 dependencies = [
  "bytemuck",
  "safe_arch",
 ]
 
-[[package]]
-name = "widestring"
-version = "1.1.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "7219d36b6eac893fa81e84ebe06485e7dcbb616177469b142df14f1f4deb1311"
-
-[[package]]
-name = "winapi"
-version = "0.3.9"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "5c839a674fcd7a98952e593242ea400abe93992746761e38641405d28b00f419"
-dependencies = [
- "winapi-i686-pc-windows-gnu",
- "winapi-x86_64-pc-windows-gnu",
-]
-
-[[package]]
-name = "winapi-i686-pc-windows-gnu"
-version = "0.4.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "ac3b87c63620426dd9b991e5ce0329eff545bccbbb34f3be09ff6fb6ab51b7b6"
-
 [[package]]
 name = "winapi-util"
 version = "0.1.9"
@@ -2035,17 +2020,11 @@ dependencies = [
  "windows-sys 0.59.0",
 ]
 
-[[package]]
-name = "winapi-x86_64-pc-windows-gnu"
-version = "0.4.0"
-source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "712e227841d057c1ee1cd2fb22fa7e5a5461ae8e48fa2ca79ec42cfc1931183f"
-
 [[package]]
 name = "windows"
-version = "0.52.0"
+version = "0.58.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "e48a53791691ab099e5e2ad123536d0fff50652600abaf43bbf952894110d0be"
+checksum = "dd04d41d93c4992d421894c18c8b43496aa748dd4c081bac0dc93eb0489272b6"
 dependencies = [
  "windows-core",
  "windows-targets 0.52.6",
@@ -2053,10 +2032,55 @@ dependencies = [
 
 [[package]]
 name = "windows-core"
-version = "0.52.0"
+version = "0.58.0"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "33ab640c8d7e35bf8ba19b884ba838ceb4fba93a4e8c65a9059d08afcfc683d9"
+checksum = "6ba6d44ec8c2591c134257ce647b7ea6b20335bf6379a27dac5f1641fcf59f99"
 dependencies = [
+ "windows-implement",
+ "windows-interface",
+ "windows-result",
+ "windows-strings",
+ "windows-targets 0.52.6",
+]
+
+[[package]]
+name = "windows-implement"
+version = "0.58.0"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "2bbd5b46c938e506ecbce286b6628a02171d56153ba733b6c741fc627ec9579b"
+dependencies = [
+ "proc-macro2",
+ "quote",
+ "syn",
+]
+
+[[package]]
+name = "windows-interface"
+version = "0.58.0"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "053c4c462dc91d3b1504c6fe5a726dd15e216ba718e84a0e46a88fbe5ded3515"
+dependencies = [
+ "proc-macro2",
+ "quote",
+ "syn",
+]
+
+[[package]]
+name = "windows-result"
+version = "0.2.0"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "1d1043d8214f791817bab27572aaa8af63732e11bf84aa21a45a78d6c317ae0e"
+dependencies = [
+ "windows-targets 0.52.6",
+]
+
+[[package]]
+name = "windows-strings"
+version = "0.1.0"
+source = "registry+https://github.com/rust-lang/crates.io-index"
+checksum = "4cd9b125c486025df0eabcb585e62173c6c9eddcec5d117d3b6e8c30e2ee4d10"
+dependencies = [
+ "windows-result",
  "windows-targets 0.52.6",
 ]
 
@@ -2267,18 +2291,18 @@ checksum = "589f6da84c646204747d1270a2a5661ea66ed1cced2631d546fdfb155959f9ec"
 
 [[package]]
 name = "winit"
-version = "0.30.5"
+version = "0.30.9"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "0be9e76a1f1077e04a411f0b989cbd3c93339e1771cb41e71ac4aee95bfd2c67"
+checksum = "a809eacf18c8eca8b6635091543f02a5a06ddf3dad846398795460e6e0ae3cc0"
 dependencies = [
  "ahash",
  "android-activity",
  "atomic-waker",
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "block2",
  "bytemuck",
  "calloop",
- "cfg_aliases 0.2.1",
+ "cfg_aliases",
  "concurrent-queue",
  "core-foundation",
  "core-graphics",
@@ -2319,9 +2343,9 @@ dependencies = [
 
 [[package]]
 name = "winnow"
-version = "0.6.20"
+version = "0.7.4"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "36c1fec1a2bb5866f07c25f68c26e565c4c200aebb96d7e55710c19d3e8ac49b"
+checksum = "0e97b544156e9bebe1a0ffbc03484fc1ffe3100cbce3ffb17eac35f7cdd7ab36"
 dependencies = [
  "memchr",
 ]
@@ -2370,7 +2394,7 @@ version = "0.4.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "d039de8032a9a8856a6be89cea3e5d12fdd82306ab7c94d74e6deab2460651c5"
 dependencies = [
- "bitflags 2.6.0",
+ "bitflags 2.9.0",
  "dlib",
  "log",
  "once_cell",
@@ -2385,9 +2409,9 @@ checksum = "b9cc00251562a284751c9973bace760d86c0276c471b4be569fe6b068ee97a56"
 
 [[package]]
 name = "xml-rs"
-version = "0.8.22"
+version = "0.8.25"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "af4e2e2f7cba5a093896c1e150fbfe177d1883e7448200efb81d40b9d339ef26"
+checksum = "c5b940ebc25896e71dd073bad2dbaa2abfe97b0a391415e22ad1326d9c54e3c4"
 
 [[package]]
 name = "zerocopy"
@@ -2406,5 +2430,5 @@ checksum = "fa4f8080344d4671fb4e831a13ad1e68092748387dfc4f55e356242fae12ce3e"
 dependencies = [
  "proc-macro2",
  "quote",
- "syn 2.0.84",
+ "syn",
 ]
diff --git a/Cargo.toml b/Cargo.toml
index a8f27a7..40d8fe3 100644
--- a/Cargo.toml
+++ b/Cargo.toml
@@ -4,9 +4,9 @@ version = "0.1.0"
 edition = "2021"
 
 [dependencies]
-winit = "0.30.5"
-wgpu = "22.1.0"
-nalgebra = "0.33.1"
+winit = "0.30.9"
 pollster = "0.3.0"
 bytemuck = "1.19.0"
 num-traits = "0.2.19"
+nalgebra = "0.33.2"
+wgpu = "24.0.1"
diff --git a/src/client/camera.rs b/src/client/camera.rs
index 6a4a50f..acfb4fb 100644
--- a/src/client/camera.rs
+++ b/src/client/camera.rs
@@ -50,6 +50,9 @@ impl Zoom {
     pub fn level(&self) -> i32 {
         self.level
     }
+    pub fn rel_zoom(&self) -> f32 {
+        self.exp.exp2()
+    }
 }
 
 impl AddAssign<f32> for Zoom {
diff --git a/src/client/handle_input.rs b/src/client/handle_input.rs
index 34ba234..3f71ff2 100644
--- a/src/client/handle_input.rs
+++ b/src/client/handle_input.rs
@@ -16,7 +16,6 @@ impl Client<'_> {
         }
 
         let speed = FixedDec::from(per_sec * 0.5) * camera.zoom.mult();
-        let old = f32::from(&camera.pos.x);
         if input.pressed(K::KeyW) {
             camera.pos.y += &speed;
         }
@@ -29,9 +28,5 @@ impl Client<'_> {
         if input.pressed(K::KeyD) {
             camera.pos.x += &speed;
         }
-        let new = f32::from(&camera.pos.x);
-        if (new - old).abs() > 0.5 {
-            println!("{} + {} = {}", old, f32::from(speed), new);
-        }
     }
 }
diff --git a/src/client/mod.rs b/src/client/mod.rs
index 5d7e3a1..f55f143 100644
--- a/src/client/mod.rs
+++ b/src/client/mod.rs
@@ -60,8 +60,7 @@ impl Client<'_> {
             WindowEvent::CloseRequested => self.exit = true,
             WindowEvent::Resized(size) => self.renderer.resize(size),
             WindowEvent::RedrawRequested => {
-                self.renderer.update(&self.camera);
-                self.renderer.draw();
+                self.renderer.render(&self.camera);
                 self.window.request_redraw();
             }
             WindowEvent::CursorLeft { .. } => {
diff --git a/src/client/render/tile/data.rs b/src/client/render/compute/data.rs
similarity index 57%
rename from src/client/render/tile/data.rs
rename to src/client/render/compute/data.rs
index 6ea9c93..025657f 100644
--- a/src/client/render/tile/data.rs
+++ b/src/client/render/compute/data.rs
@@ -6,28 +6,36 @@ use super::Camera;
 
 const VIEW_ALIGN: usize = 4 * 2;
 
-pub struct View {
+pub struct ComputeView {
     pub bytes: Vec<u8>,
 }
 
-impl View {
+impl ComputeView {
     pub fn bytes(&self) -> &[u8] {
         &self.bytes
     }
 }
 
-impl Default for View {
+impl Default for ComputeView {
     fn default() -> Self {
         let val = FixedDec::from_parts(false, 0, vec![0, 0, 0]);
-        Self::new(Vector2::zeros(), 0, &val, &val, &val)
+        Self::new(true, Vector2::zeros(), 0, &val, &val, &val)
     }
 }
 
-impl View {
-    fn new(stretch: Vector2<f32>, level: i32, scale: &FixedDec, x: &FixedDec, y: &FixedDec) -> Self {
+impl ComputeView {
+    fn new(
+        reset: bool,
+        dims: Vector2<u32>,
+        level: i32,
+        scale: &FixedDec,
+        x: &FixedDec,
+        y: &FixedDec,
+    ) -> Self {
         let mut bytes = Vec::new();
-        bytes.extend(bytemuck::cast_slice(&[stretch.x, stretch.y]));
+        bytes.extend((reset as u32).to_le_bytes());
         bytes.extend(level.to_le_bytes());
+        bytes.extend(bytemuck::cast_slice(&[dims.x, dims.y]));
         scale.to_bytes(&mut bytes);
         x.to_bytes(&mut bytes);
         y.to_bytes(&mut bytes);
@@ -35,16 +43,16 @@ impl View {
         if rem != 0 {
             bytes.extend((0..(VIEW_ALIGN - rem)).map(|_| 0));
         }
-        Self{ bytes }
+        Self { bytes }
     }
 
-    pub fn from_camera_size(camera: &Camera, size: &Vector2<u32>) -> Self {
+    pub fn from_camera_size(camera: &Camera, size: &Vector2<u32>, reset: bool, len: usize) -> Self {
         let mut x = camera.pos.x.clone();
         x.set_whole_len(1);
-        x.set_dec_len(2);
+        x.set_dec_len(len as i32 - 1);
         let mut y = camera.pos.y.clone();
         y.set_whole_len(1);
-        y.set_dec_len(2);
+        y.set_dec_len(len as i32 - 1);
 
         let fsize: Vector2<f32> = size.cast();
         let stretch = if size.x < size.y {
@@ -54,8 +62,14 @@ impl View {
         };
 
         let mut scale = camera.zoom.mult().clone();
-        scale.set_precision(3);
+        scale.set_precision(len);
 
-        Self::new(stretch, camera.zoom.level(), &scale, &x, &y)
+        Self::new(reset, *size, camera.zoom.level(), &scale, &x, &y)
+    }
+}
+
+impl PartialEq for ComputeView {
+    fn eq(&self, other: &Self) -> bool {
+        self.bytes[1..] == other.bytes[1..]
     }
 }
diff --git a/src/client/render/tile/render.wgsl b/src/client/render/compute/fixed.wgsl
similarity index 78%
rename from src/client/render/tile/render.wgsl
rename to src/client/render/compute/fixed.wgsl
index 9d57402..fdc6a87 100644
--- a/src/client/render/tile/render.wgsl
+++ b/src/client/render/compute/fixed.wgsl
@@ -1,14 +1,5 @@
-const LEN: u32 = 3;
-const ILEN: i32 = 3;
-const LEN2: u32 = LEN * 2;
-
-struct View {
-    stretch: vec2<f32>,
-    level: i32,
-    scale: FixedDec,
-    x: FixedDec,
-    y: FixedDec,
-}
+const POS: u32 = 0u;
+const NEG: u32 = 1u;
 
 struct FixedDec {
     sign: u32,
@@ -16,63 +7,18 @@ struct FixedDec {
     parts: array<u32, LEN>,
 }
 
-@group(0) @binding(0)
-var<storage> view: View;
-
-struct VertexOutput {
-    @builtin(position) vertex_pos: vec4<f32>,
-    @location(0) world_pos: vec2<f32>,
-};
-
-@vertex
-fn vs_main(
-    @builtin(vertex_index) vi: u32,
-    @builtin(instance_index) ii: u32,
-) -> VertexOutput {
-    var out: VertexOutput;
-
-    let pos = vec2<f32>(
-        f32(vi % 2u),
-        f32(vi / 2u),
-    ) * 2.0 - 1.0;
-    out.vertex_pos = vec4<f32>(pos.x, -pos.y, 0.0, 1.0);
-    out.world_pos = pos;
-    out.world_pos.y *= -1.0;
-    out.world_pos *= view.stretch;
-    return out;
-}
-
-@fragment
-fn fs_main(
-    in: VertexOutput,
-) -> @location(0) vec4<f32> {
-    let cx = add(mul(from_f32(in.world_pos.x), view.scale), view.x);
-    let cy = add(mul(from_f32(in.world_pos.y), view.scale), view.y);
-    var x = zero();
-    var y = zero();
-    let two = from_f32(2.0);
-    let thresh = from_f32(2.0 * 2.0);
-    var i = 0u;
-    let max = 50u + (1u << u32(view.level));
-    loop {
-        let x2 = mul(x, x);
-        let y2 = mul(y, y);
-        if gt(add(x2, y2), thresh) || i >= max {
-            break;
-        }
-        y = add(mul(two, mul(x, y)), cy);
-        x = add(sub(x2, y2), cx);
-        i += 1u;
+fn shr(lhs: FixedDec, rhs: i32) -> FixedDec {
+    var parts = array<u32, LEN>();
+    let sr = u32(rem_euclid(rhs, 32));
+    let sl = 32 - sr;
+    let mask = (1u << sr) - 1;
+    var rem = 0u;
+    for (var i = 0; i < ILEN; i += 1) {
+        let part = lhs.parts[i];
+        parts[i] = (part >> sr) ^ rem;
+        rem = (part & mask) << sl;
     }
-    var color = vec3<f32>(0.0, 0.0, 0.0);
-    if i != max {
-        let pi = 3.1415;
-        let hue = f32(i) / 30.0;
-        color.r = cos(hue);
-        color.g = cos(hue - 2.0 * pi / 3.0);
-        color.b = cos(hue - 4.0 * pi / 3.0);
-    }
-    return vec4(color, 1.0);
+    return FixedDec(lhs.sign, lhs.dec, parts);
 }
 
 fn add(lhs: FixedDec, rhs: FixedDec) -> FixedDec {
@@ -136,9 +82,6 @@ fn at(dec: FixedDec, i: i32) -> u32 {
     return parts[i];
 }
 
-const POS: u32 = 0u;
-const NEG: u32 = 1u;
-
 fn mul(lhs: FixedDec, rhs: FixedDec) -> FixedDec {
     let sign = u32(lhs.sign != rhs.sign);
     var parts = array<u32, LEN2>();
@@ -173,7 +116,7 @@ fn mul(lhs: FixedDec, rhs: FixedDec) -> FixedDec {
             let res2 = res + carry;
             let carry2 = res2 < res;
             parts[k] = res2;
-            carry = u32(carry1 || carry2) + msb;
+            carry = u32(carry1) + u32(carry2) + msb;
         }
         parts[i] = carry;
     }
@@ -202,6 +145,25 @@ fn gt(x: FixedDec, y: FixedDec) -> bool {
     return x.parts[0] > y.parts[0];
 }
 
+fn eq(x: FixedDec, y: FixedDec) -> bool {
+    if x.sign != y.sign || x.dec != y.dec {
+        return false;
+    }
+    var i = 0u;
+    var xp = x.parts;
+    var yp = y.parts;
+    while i < LEN - 2 {
+        if xp[i] != yp[i] {
+            return false;
+        }
+        i += 1u;
+    }
+    if abs(f32(xp[i]) - f32(yp[i])) > 1024 * 16 {
+        return false;
+    }
+    return true;
+}
+
 fn to_f32(value: FixedDec) -> f32 {
     var parts = value.parts;
 
diff --git a/src/client/render/compute/layout.rs b/src/client/render/compute/layout.rs
new file mode 100644
index 0000000..4190fed
--- /dev/null
+++ b/src/client/render/compute/layout.rs
@@ -0,0 +1,123 @@
+use wgpu::{PipelineCompilationOptions, ShaderStages};
+
+use crate::client::render::util::ArrayBuffer;
+
+use super::{
+    util::{Storage, Texture},
+    ComputeView,
+};
+
+pub struct Layout {
+    bind_layout: wgpu::BindGroupLayout,
+    pipeline_layout: wgpu::PipelineLayout,
+    format: wgpu::TextureFormat,
+    pub output: Texture,
+    pub view: Storage,
+    pub work: ArrayBuffer<u32>,
+}
+
+impl Layout {
+    pub fn init(device: &wgpu::Device, config: &wgpu::SurfaceConfiguration, len: usize) -> Self {
+        let view = Storage::init_with(device, "view", ComputeView::default().bytes());
+        let work = ArrayBuffer::init_with(
+            device,
+            "test",
+            wgpu::BufferUsages::STORAGE,
+            &work_vec(config.width, config.height, len),
+        );
+
+        let desc = wgpu::TextureDescriptor {
+            label: Some("compute output"),
+            size: wgpu::Extent3d {
+                width: config.width,
+                height: config.height,
+                depth_or_array_layers: 1,
+            },
+            mip_level_count: 1,
+            sample_count: 1,
+            dimension: wgpu::TextureDimension::D2,
+            format: wgpu::TextureFormat::Rgba8Unorm,
+            usage: wgpu::TextureUsages::STORAGE_BINDING | wgpu::TextureUsages::TEXTURE_BINDING,
+            view_formats: &[],
+        };
+        let output = Texture::init(
+            device,
+            desc,
+            wgpu::TextureViewDescriptor::default(),
+            wgpu::SamplerDescriptor::default(),
+        );
+
+        let bind_layout = device.create_bind_group_layout(&wgpu::BindGroupLayoutDescriptor {
+            entries: &[
+                view.bind_group_layout_entry(0, true, wgpu::ShaderStages::COMPUTE),
+                work.bind_group_layout_entry(
+                    1,
+                    wgpu::BufferBindingType::Storage { read_only: false },
+                    wgpu::ShaderStages::COMPUTE,
+                ),
+                wgpu::BindGroupLayoutEntry {
+                    binding: 2,
+                    visibility: ShaderStages::COMPUTE,
+                    ty: wgpu::BindingType::StorageTexture {
+                        access: wgpu::StorageTextureAccess::WriteOnly,
+                        format: output.format(),
+                        view_dimension: wgpu::TextureViewDimension::D2,
+                    },
+                    count: None,
+                },
+            ],
+            label: Some("compute"),
+        });
+
+        let pipeline_layout = device.create_pipeline_layout(&wgpu::PipelineLayoutDescriptor {
+            label: Some("Tile Pipeline Layout"),
+            bind_group_layouts: &[&bind_layout],
+            push_constant_ranges: &[],
+        });
+
+        Self {
+            view,
+            output,
+            bind_layout,
+            pipeline_layout,
+            work,
+            format: config.format,
+        }
+    }
+
+    pub fn bind_group(&self, device: &wgpu::Device) -> wgpu::BindGroup {
+        device.create_bind_group(&wgpu::BindGroupDescriptor {
+            layout: &self.bind_layout,
+            entries: &[
+                self.view.bind_group_entry(0),
+                self.work.bind_group_entry(1),
+                self.output.view_bind_group_entry(2),
+            ],
+            label: Some("voxel render"),
+        })
+    }
+
+    pub fn pipeline(
+        &self,
+        device: &wgpu::Device,
+        shader: &wgpu::ShaderModule,
+    ) -> wgpu::ComputePipeline {
+        device.create_compute_pipeline(&wgpu::ComputePipelineDescriptor {
+            label: Some("Voxel Pipeline"),
+            layout: Some(&self.pipeline_layout),
+            entry_point: Some("main"),
+            module: shader,
+            cache: None,
+            compilation_options: PipelineCompilationOptions::default(),
+        })
+    }
+}
+
+pub fn work_size(width: u32, height: u32, len: usize) -> usize {
+    let varwidth = (2 + len) * 2;
+    (width * height) as usize * (varwidth * 2 + 1)
+}
+
+pub fn work_vec(width: u32, height: u32, len: usize) -> Vec<u32> {
+    vec![0u32; work_size(width, height, len)]
+}
diff --git a/src/client/render/compute/mod.rs b/src/client/render/compute/mod.rs
new file mode 100644
index 0000000..1582ad0
--- /dev/null
+++ b/src/client/render/compute/mod.rs
@@ -0,0 +1,110 @@
+use std::ops::{Deref, DerefMut};
+
+mod data;
+mod layout;
+
+use super::*;
+pub use data::*;
+use layout::*;
+
+pub struct ComputePipeline {
+    layout: Layout,
+    pipeline: wgpu::ComputePipeline,
+    bind_group: wgpu::BindGroup,
+    old_view: ComputeView,
+    old_len: usize,
+}
+
+const FIXED_SHADER: &str = include_str!("fixed.wgsl");
+const SHADER: &str = include_str!("shader.wgsl");
+
+impl ComputePipeline {
+    pub fn init(device: &wgpu::Device, config: &wgpu::SurfaceConfiguration, len: usize) -> Self {
+        let layout = Layout::init(device, config, len);
+        Self {
+            pipeline: layout.pipeline(device, &Self::shader(device, len)),
+            bind_group: layout.bind_group(device),
+            layout,
+            old_view: ComputeView::default(),
+            old_len: len,
+        }
+    }
+
+    pub fn update(
+        &mut self,
+        device: &wgpu::Device,
+        encoder: &mut wgpu::CommandEncoder,
+        belt: &mut wgpu::util::StagingBelt,
+        camera: &Camera,
+        size: &Vector2<u32>,
+        len: usize,
+    ) {
+        let mut view = ComputeView::from_camera_size(camera, size, false, len);
+        if view != self.old_view {
+            for (i, b) in 1u32.to_le_bytes().iter().enumerate() {
+                view.bytes[i] = *b;
+            }
+        }
+        if len != self.old_len {
+            self.old_len = len;
+            self.pipeline = self.pipeline(device, &Self::shader(device, len));
+            self.work.set(work_vec(size.x, size.y, len));
+        }
+        let updated = self.work.update(device, encoder, belt)
+            | self.view.update(device, encoder, belt, view.bytes());
+        if updated {
+            self.bind_group = self.layout.bind_group(device);
+        }
+        self.old_view = view;
+    }
+
+    pub fn run(&self, encoder: &mut wgpu::CommandEncoder) {
+        let mut pass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor::default());
+        pass.set_pipeline(&self.pipeline);
+        pass.set_bind_group(0, &self.bind_group, &[]);
+        pass.dispatch_workgroups(240, 135, 1);
+    }
+
+    pub fn resize(
+        &mut self,
+        device: &wgpu::Device,
+        encoder: &mut wgpu::CommandEncoder,
+        belt: &mut wgpu::util::StagingBelt,
+        size: Vector2<u32>,
+        len: usize,
+    ) {
+        self.work.set(work_vec(size.x, size.y, len));
+        self.old_len = len;
+        self.output.resize(
+            device,
+            wgpu::Extent3d {
+                width: size.x,
+                height: size.y,
+                depth_or_array_layers: 1,
+            },
+        );
+        self.bind_group = self.layout.bind_group(device);
+    }
+
+    pub fn shader(device: &wgpu::Device, len: usize) -> wgpu::ShaderModule {
+        let string = FIXED_SHADER.to_string() + &SHADER.replace("REPLACE_LEN", &format!("{}", len));
+        device.create_shader_module(wgpu::ShaderModuleDescriptor {
+            label: Some("compute"),
+            source: wgpu::ShaderSource::Wgsl(string.into()),
+        })
+    }
+}
+
+impl Deref for ComputePipeline {
+    type Target = Layout;
+
+    fn deref(&self) -> &Self::Target {
+        &self.layout
+    }
+}
+
+impl DerefMut for ComputePipeline {
+    fn deref_mut(&mut self) -> &mut Self::Target {
+        &mut self.layout
+    }
+}
diff --git a/src/client/render/compute/shader.wgsl b/src/client/render/compute/shader.wgsl
new file mode 100644
index 0000000..99fb32f
--- /dev/null
+++ b/src/client/render/compute/shader.wgsl
@@ -0,0 +1,102 @@
+override CHUNK_POW: u32 = 10;
+const LEN: u32 = REPLACE_LENu;
+const ILEN: i32 = i32(LEN);
+const LEN2: u32 = LEN * 2;
+override WGX: u32 = 8;
+override WGY: u32 = 8;
+
+struct View {
+    reset: u32,
+    level: i32,
+    dims: vec2<u32>,
+    scale: FixedDec,
+    corner_x: FixedDec,
+    corner_y: FixedDec,
+}
+
+@group(0) @binding(0)
+var<storage> view: View;
+@group(0) @binding(1)
+var<storage, read_write> work: array<u32>;
+@group(0) @binding(2)
+var output: texture_storage_2d<rgba8unorm, write>;
+
+@compute @workgroup_size(WGX, WGY, 1)
+fn main(
+    @builtin(global_invocation_id) id: vec3<u32>
+) {
+    if id.x > view.dims.x - 1 || id.y > view.dims.y - 1 {
+        return;
+    }
+    // TODO: actually use width
+    let varwidth = LEN + 2;
+    let workwidth = varwidth * 2 + 1;
+    let worki = (id.x * view.dims.y + id.y) * workwidth;
+    let xidx = worki + 1;
+    let yidx = xidx + varwidth;
+
+    // let dec = view.corner_x.dec;
+    // var rel_x = FixedDec(POS, dec, array<u32, LEN>());
+    // rel_x.parts[0] = id.x;
+    // rel_x = shr(rel_x, view.level);
+    // var rel_y = FixedDec(POS, dec, array<u32, LEN>());
+    // rel_y.parts[0] = id.y;
+    // rel_y = shr(rel_y, view.level);
+    // let cx = add(view.corner_x, rel_x);
+    // let cy = add(view.corner_y, rel_y);
+    let fdims = vec2<f32>(view.dims);
+    var stretch: vec2<f32>;
+    if fdims.x < fdims.y {
+        stretch = vec2(fdims.x / fdims.y, 1.0);
+    } else {
+        stretch = vec2(1.0, fdims.y / fdims.x);
+    }
+    let fpos = (vec2<f32>(id.xy) / fdims - 0.5) * stretch;
+    let cx = add(mul(from_f32(fpos.x), view.scale), view.corner_x);
+    let cy = add(mul(from_f32(fpos.y), view.scale), view.corner_y);
+    var x: FixedDec;
+    var y: FixedDec;
+    var i = work[worki];
+    if bool(view.reset) {
+        x = zero();
+        y = zero();
+        i = 0;
+    } else {
+        x = FixedDec(work[xidx + 0], bitcast<i32>(work[xidx + 1]), array<u32, LEN>());
+        y = FixedDec(work[yidx + 0], bitcast<i32>(work[yidx + 1]), array<u32, LEN>());
+        for (var j = 0u; j < LEN; j += 1u) {
+            x.parts[j] = work[xidx + 2 + j];
+            y.parts[j] = work[yidx + 2 + j];
+        }
+    }
+    let max = i + 1;
+    let thresh = from_f32(2.0 * 2.0);
+    loop {
+        let x2 = mul(x, x);
+        let y2 = mul(y, y);
+        if gt(add(x2, y2), thresh) || i >= max {
+            break;
+        }
+        let xy = mul(x, y);
+        y = add(add(xy, xy), cy);
+        x = add(sub(x2, y2), cx);
+        i += 1u;
+    }
+    work[worki] = i;
+    work[xidx + 0] = x.sign; work[xidx + 1] = bitcast<u32>(x.dec);
+    work[yidx + 0] = y.sign; work[yidx + 1] = bitcast<u32>(y.dec);
+    for (var j = 0u; j < LEN; j += 1u) {
+        work[xidx + 2 + j] = x.parts[j];
+        work[yidx + 2 + j] = y.parts[j];
+    }
+    var color = vec3<f32>(0.0, 0.0, 0.0);
+    if i != max {
+        let pi = 3.1415;
+        let hue = f32(i) / 30.0;
+        color.r = cos(hue);
+        color.g = cos(hue - 2.0 * pi / 3.0);
+        color.b = cos(hue - 4.0 * pi / 3.0);
+    }
+    textureStore(output, id.xy, vec4(color, 1.0));
+}
+
diff --git a/src/client/render/mod.rs b/src/client/render/mod.rs
index 0d2a2a7..2adc02f 100644
--- a/src/client/render/mod.rs
+++ b/src/client/render/mod.rs
@@ -1,15 +1,22 @@
-mod tile;
+mod compute;
+mod output;
 mod util;
+mod view;
 
 use std::sync::Arc;
 
+use compute::ComputePipeline;
 use nalgebra::Vector2;
-use tile::TilePipeline;
+use output::RenderPipeline;
 use util::GPUTimer;
+use view::ChunkView;
 use winit::{dpi::PhysicalSize, window::Window};
 
 use super::camera::Camera;
 
+const CHUNK_POW: u32 = 7;
+const CHUNK_WIDTH: u32 = 2u32.pow(CHUNK_POW);
+
 pub struct Renderer<'a> {
     size: Vector2<u32>,
     surface: wgpu::Surface<'a>,
@@ -19,15 +26,18 @@ pub struct Renderer<'a> {
     config: wgpu::SurfaceConfiguration,
     staging_belt: wgpu::util::StagingBelt,
     timer: GPUTimer,
+    chunk_view: ChunkView,
+    len: usize,
 
-    tile_pipeline: TilePipeline,
+    compute_pipeline: ComputePipeline,
+    render_pipeline: RenderPipeline,
 }
 
 impl Renderer<'_> {
     pub fn new(window: Arc<Window>) -> Self {
         let size = window.inner_size();
 
-        let instance = wgpu::Instance::new(wgpu::InstanceDescriptor {
+        let instance = wgpu::Instance::new(&wgpu::InstanceDescriptor {
             backends: wgpu::Backends::PRIMARY,
             ..Default::default()
         });
@@ -90,8 +100,14 @@ impl Renderer<'_> {
         let staging_belt = wgpu::util::StagingBelt::new(1024);
         let timer = GPUTimer::new(&device, queue.get_timestamp_period(), 1);
 
+        let len = 2;
+
+        let compute_pipeline = ComputePipeline::init(&device, &config, len);
+        let render_pipeline = RenderPipeline::init(&device, &config, &compute_pipeline.output);
+
         Self {
-            tile_pipeline: TilePipeline::init(&device, &config),
+            render_pipeline,
+            compute_pipeline,
             size: Vector2::new(size.width, size.height),
             staging_belt,
             surface,
@@ -100,47 +116,44 @@ impl Renderer<'_> {
             device,
             config,
             queue,
+            chunk_view: ChunkView::new(),
+            len,
         }
     }
 
-    pub fn update(&mut self, camera: &Camera) {
-        self.tile_pipeline.update(
+    pub fn render(&mut self, camera: &Camera) {
+
+        self.len = (camera.zoom.level() as f32 / 15.0 + 2.0).round() as usize;
+        println!("{}", self.len);
+        // let new = (camera.zoom.level() as f32 / 15.0 + 2.0).round() as usize;
+        // println!("{}", new);
+
+        self.compute_pipeline.update(
             &self.device,
             &mut self.encoder,
             &mut self.staging_belt,
             camera,
             &self.size,
+            self.len,
+        );
+        self.chunk_view.update(camera, &self.size);
+        self.render_pipeline.update(
+            &self.device,
+            &mut self.encoder,
+            &mut self.staging_belt,
+            &self.chunk_view.render,
         );
-    }
 
-    pub fn draw(&mut self) {
         let mut encoder = std::mem::replace(&mut self.encoder, Self::create_encoder(&self.device));
         let output = self.surface.get_current_texture().unwrap();
-        let view = output
-            .texture
-            .create_view(&wgpu::TextureViewDescriptor::default());
 
         self.timer.start(&mut encoder, 0);
-        let mut render_pass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor {
-            label: Some("Render Pass"),
-            color_attachments: &[Some(wgpu::RenderPassColorAttachment {
-                view: &view,
-                resolve_target: None,
-                ops: wgpu::Operations {
-                    load: wgpu::LoadOp::Clear(wgpu::Color::BLACK),
-                    store: wgpu::StoreOp::Store,
-                },
-            })],
-            depth_stencil_attachment: None,
-            timestamp_writes: None,
-            occlusion_query_set: None,
-        });
-        self.tile_pipeline.draw(&mut render_pass);
-        drop(render_pass);
+        self.compute_pipeline.run(&mut encoder);
         self.timer.stop(&mut encoder, 0);
-
         self.timer.resolve(&mut encoder);
 
+        self.render_pipeline.draw(&mut encoder, &output);
+
         self.staging_belt.finish();
         self.queue.submit(std::iter::once(encoder.finish()));
         output.present();
@@ -154,6 +167,8 @@ impl Renderer<'_> {
         self.config.width = size.width;
         self.config.height = size.height;
         self.surface.configure(&self.device, &self.config);
+        self.compute_pipeline.resize(&self.device, &mut self.encoder, &mut self.staging_belt, self.size, self.len);
+        self.render_pipeline.resize(&self.device, self.size, &self.compute_pipeline.output);
     }
 
     fn create_encoder(device: &wgpu::Device) -> wgpu::CommandEncoder {
diff --git a/src/client/render/output/data.rs b/src/client/render/output/data.rs
new file mode 100644
index 0000000..cc85c05
--- /dev/null
+++ b/src/client/render/output/data.rs
@@ -0,0 +1,45 @@
+use nalgebra::Vector2;
+
+use crate::client::render::CHUNK_POW;
+
+use super::{Camera, CHUNK_WIDTH};
+
+#[repr(C, align(8))]
+#[derive(Clone, Copy, Default, PartialEq)]
+pub struct WindowView {
+    pub stretch: Vector2<f32>,
+    pub pos: Vector2<f32>,
+    pub rendered_chunks: Vector2<u32>,
+}
+
+unsafe impl bytemuck::Pod for WindowView {}
+unsafe impl bytemuck::Zeroable for WindowView {}
+
+impl WindowView {
+    pub fn from_camera_size(camera: &Camera, size: &Vector2<u32>) -> Self {
+        let visible_chunks = (size * 2 / CHUNK_WIDTH).add_scalar(1);
+        let rendered_chunks = Vector2::new(
+            visible_chunks.x.next_power_of_two(),
+            visible_chunks.y.next_power_of_two(),
+        );
+        let adj_zoom = camera.zoom.level() - CHUNK_POW as i32;
+        let pos = camera.pos.zip_map(&rendered_chunks, |pos, rc| {
+            let p = (pos << adj_zoom).with_lens(1, 1);
+            let (pw, pd) = p.split_whole_dec();
+            let mut chunk = (pw.parts().first().unwrap_or(&0) & (rc - 1)) as f32;
+            if pw.is_neg() {
+                chunk = rc as f32 - chunk;
+            }
+            let dec = f32::from(pd);
+            chunk + dec
+        });
+
+        let stretch = size.cast::<f32>() * camera.zoom.rel_zoom() / (CHUNK_WIDTH as f32);
+
+        Self {
+            pos,
+            stretch,
+            rendered_chunks,
+        }
+    }
+}
diff --git a/src/client/render/output/layout.rs b/src/client/render/output/layout.rs
new file mode 100644
index 0000000..f23d310
--- /dev/null
+++ b/src/client/render/output/layout.rs
@@ -0,0 +1,159 @@
+use nalgebra::Vector2;
+
+use crate::client::render::util::Texture;
+
+use super::{
+    util::{ResizableTexture, Storage},
+    WindowView, CHUNK_WIDTH,
+};
+
+pub struct Layout {
+    render_bind_layout: wgpu::BindGroupLayout,
+    render_pipeline_layout: wgpu::PipelineLayout,
+    format: wgpu::TextureFormat,
+    pub view: Storage,
+    pub chunks: ResizableTexture,
+}
+
+pub const LABEL: &str = file!();
+
+impl Layout {
+    pub fn init(
+        device: &wgpu::Device,
+        config: &wgpu::SurfaceConfiguration,
+        input: &Texture,
+    ) -> Self {
+        let view = Storage::init_with(device, "view", bytemuck::bytes_of(&WindowView::default()));
+
+        let texture_desc = wgpu::TextureDescriptor {
+            label: Some("chunks"),
+            size: wgpu::Extent3d {
+                width: CHUNK_WIDTH,
+                height: CHUNK_WIDTH,
+                depth_or_array_layers: 1,
+            },
+            mip_level_count: 1,
+            sample_count: 1,
+            dimension: wgpu::TextureDimension::D2,
+            format: wgpu::TextureFormat::R32Uint,
+            usage: wgpu::TextureUsages::TEXTURE_BINDING | wgpu::TextureUsages::STORAGE_BINDING,
+            view_formats: &[wgpu::TextureFormat::R32Uint],
+        };
+        let view_desc = wgpu::TextureViewDescriptor {
+            label: Some("chunk view"),
+            dimension: Some(wgpu::TextureViewDimension::D2Array),
+            ..Default::default()
+        };
+        let chunks = ResizableTexture::new(device, texture_desc, view_desc);
+
+        let render_bind_layout =
+            device.create_bind_group_layout(&wgpu::BindGroupLayoutDescriptor {
+                entries: &[
+                    view.bind_group_layout_entry(
+                        0,
+                        true,
+                        wgpu::ShaderStages::FRAGMENT | wgpu::ShaderStages::VERTEX,
+                    ),
+                    wgpu::BindGroupLayoutEntry {
+                        binding: 1,
+                        visibility: wgpu::ShaderStages::FRAGMENT,
+                        ty: wgpu::BindingType::Texture {
+                            sample_type: wgpu::TextureSampleType::Uint,
+                            view_dimension: wgpu::TextureViewDimension::D2Array,
+                            multisampled: false,
+                        },
+                        count: None,
+                    },
+                    wgpu::BindGroupLayoutEntry {
+                        binding: 2,
+                        visibility: wgpu::ShaderStages::FRAGMENT | wgpu::ShaderStages::VERTEX,
+                        ty: wgpu::BindingType::Texture {
+                            multisampled: false,
+                            view_dimension: wgpu::TextureViewDimension::D2,
+                            sample_type: wgpu::TextureSampleType::Float { filterable: true },
+                        },
+                        count: None,
+                    },
+                    wgpu::BindGroupLayoutEntry {
+                        binding: 3,
+                        visibility: wgpu::ShaderStages::FRAGMENT,
+                        ty: wgpu::BindingType::Sampler(wgpu::SamplerBindingType::Filtering),
+                        count: None,
+                    },
+                ],
+                label: Some(LABEL),
+            });
+
+        let render_pipeline_layout =
+            device.create_pipeline_layout(&wgpu::PipelineLayoutDescriptor {
+                label: Some(LABEL),
+                bind_group_layouts: &[&render_bind_layout],
+                push_constant_ranges: &[],
+            });
+
+        Self {
+            view,
+            chunks,
+            render_bind_layout,
+            render_pipeline_layout,
+            format: config.format,
+        }
+    }
+
+    pub fn bind_group(&self, device: &wgpu::Device, input: &Texture) -> wgpu::BindGroup {
+        device.create_bind_group(&wgpu::BindGroupDescriptor {
+            layout: &self.render_bind_layout,
+            entries: &[
+                self.view.bind_group_entry(0),
+                self.chunks.view_entry(1),
+                input.view_bind_group_entry(2),
+                input.sampler_bind_group_entry(3),
+            ],
+            label: Some(LABEL),
+        })
+    }
+
+    pub fn pipeline(
+        &self,
+        device: &wgpu::Device,
+        shader: &wgpu::ShaderModule,
+    ) -> wgpu::RenderPipeline {
+        device.create_render_pipeline(&wgpu::RenderPipelineDescriptor {
+            label: Some(LABEL),
+            layout: Some(&self.render_pipeline_layout),
+            vertex: wgpu::VertexState {
+                module: shader,
+                entry_point: Some("vs_main"),
+                buffers: &[],
+                compilation_options: wgpu::PipelineCompilationOptions::default(),
+            },
+            fragment: Some(wgpu::FragmentState {
+                module: shader,
+                entry_point: Some("fs_main"),
+                targets: &[Some(wgpu::ColorTargetState {
+                    format: self.format,
+                    blend: Some(wgpu::BlendState::REPLACE),
+                    write_mask: wgpu::ColorWrites::ALL,
+                })],
+                compilation_options: wgpu::PipelineCompilationOptions::default(),
+            }),
+            primitive: wgpu::PrimitiveState {
+                topology: wgpu::PrimitiveTopology::TriangleStrip,
+                strip_index_format: None,
+                front_face: wgpu::FrontFace::Ccw,
+                cull_mode: None,
+                polygon_mode: wgpu::PolygonMode::Fill,
+                unclipped_depth: false,
+                conservative: false,
+            },
+            depth_stencil: None,
+            multisample: wgpu::MultisampleState {
+                count: 1,
+                mask: !0,
+                alpha_to_coverage_enabled: true,
+            },
+            multiview: None,
+            cache: None,
+        })
+    }
+}
diff --git a/src/client/render/output/mod.rs b/src/client/render/output/mod.rs
new file mode 100644
index 0000000..9d1bea6
--- /dev/null
+++ b/src/client/render/output/mod.rs
@@ -0,0 +1,88 @@
+use std::ops::{Deref, DerefMut};
+use wgpu::include_wgsl;
+
+mod data;
+mod layout;
+
+use super::{util::Texture, *};
+pub use data::*;
+use layout::*;
+
+pub struct RenderPipeline {
+    layout: Layout,
+    pipeline: wgpu::RenderPipeline,
+    bind_group: wgpu::BindGroup,
+    size: Vector2<u32>,
+}
+
+const SHADER: wgpu::ShaderModuleDescriptor<'_> = include_wgsl!("shader.wgsl");
+
+impl RenderPipeline {
+    pub fn init(
+        device: &wgpu::Device,
+        config: &wgpu::SurfaceConfiguration,
+        input: &Texture,
+    ) -> Self {
+        let layout = Layout::init(device, config, input);
+        let shader = device.create_shader_module(SHADER);
+        Self {
+            pipeline: layout.pipeline(device, &shader),
+            bind_group: layout.bind_group(device, input),
+            size: Vector2::zeros(),
+            layout,
+        }
+    }
+
+    pub fn update(
+        &mut self,
+        device: &wgpu::Device,
+        encoder: &mut wgpu::CommandEncoder,
+        belt: &mut wgpu::util::StagingBelt,
+        view: &WindowView,
+    ) {
+        self.view
+            .update(device, encoder, belt, bytemuck::bytes_of(view));
+    }
+
+    pub fn draw(&self, encoder: &mut wgpu::CommandEncoder, output: &wgpu::SurfaceTexture) {
+        let view = output
+            .texture
+            .create_view(&wgpu::TextureViewDescriptor::default());
+
+        let mut render_pass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor {
+            label: Some("Render Pass"),
+            color_attachments: &[Some(wgpu::RenderPassColorAttachment {
+                view: &view,
+                resolve_target: None,
+                ops: wgpu::Operations {
+                    load: wgpu::LoadOp::Clear(wgpu::Color::BLACK),
+                    store: wgpu::StoreOp::Store,
+                },
+            })],
+            depth_stencil_attachment: None,
+            timestamp_writes: None,
+            occlusion_query_set: None,
+        });
+        render_pass.set_pipeline(&self.pipeline);
+        render_pass.set_bind_group(0, &self.bind_group, &[]);
+        render_pass.draw(0..4, 0..1);
+    }
+
+    pub fn resize(&mut self, device: &wgpu::Device, size: Vector2<u32>, input: &Texture) {
+        self.bind_group = self.layout.bind_group(device, input);
+    }
+}
+
+impl Deref for RenderPipeline {
+    type Target = Layout;
+
+    fn deref(&self) -> &Self::Target {
+        &self.layout
+    }
+}
+
+impl DerefMut for RenderPipeline {
+    fn deref_mut(&mut self) -> &mut Self::Target {
+        &mut self.layout
+    }
+}
diff --git a/src/client/render/output/shader.wgsl b/src/client/render/output/shader.wgsl
new file mode 100644
index 0000000..61de6b9
--- /dev/null
+++ b/src/client/render/output/shader.wgsl
@@ -0,0 +1,76 @@
+struct View {
+    stretch: vec2<f32>,
+    pos: vec2<f32>,
+    rendered_chunks: vec2<u32>,
+}
+
+@group(0) @binding(0)
+var<storage> view: View;
+@group(0) @binding(1)
+var chunks: texture_2d_array<u32>;
+
+@group(0) @binding(2)
+var tex: texture_2d<f32>;
+@group(0) @binding(3)
+var sam: sampler;
+
+struct VertexOutput {
+    @builtin(position) vertex_pos: vec4<f32>,
+    @location(0) world_pos: vec2<f32>,
+    @location(1) tex_pos: vec2<f32>,
+};
+
+@vertex
+fn vs_main(
+    @builtin(vertex_index) vi: u32,
+    @builtin(instance_index) ii: u32,
+) -> VertexOutput {
+    var out: VertexOutput;
+
+    let pos = vec2<f32>(
+        f32(vi % 2u),
+        f32(vi / 2u),
+    ) * 2.0 - 1.0;
+    out.vertex_pos = vec4<f32>(pos.x, -pos.y, 0.0, 1.0);
+    out.world_pos = pos / 2.0;
+    out.world_pos.y *= -1.0;
+    out.world_pos *= view.stretch;
+    out.world_pos += view.pos;
+
+    let pos2 = vec2<f32>(
+        f32(vi % 2u),
+        f32(vi / 2u),
+    );
+    out.tex_pos = pos2;
+    out.tex_pos.y = 1.0 - out.tex_pos.y;
+
+    return out;
+}
+
+@fragment
+fn fs_main(
+    in: VertexOutput,
+) -> @location(0) vec4<f32> {
+    // let a = textureLoad(chunks, vec2<u32>(0), 0, 0);
+    // let rc = vec2<i32>(view.rendered_chunks);
+    // let rcf = vec2<f32>(rc);
+    // let cposi = vec2<i32>(floor(in.world_pos));
+    // let cposu = vec2<i32>(
+    //     rem_euclid(cposi.x, rc.x),
+    //     rem_euclid(cposi.y, rc.y)
+    // );
+    // let cposf = vec2<f32>(cposu);
+    // return vec4(cposf / rcf, 0.0, 1.0);
+    return textureSample(tex, sam, in.tex_pos);
+}
+
+fn div_euclid(x: i32, y: i32) -> i32 {
+    if x < 0 {
+        return -((-x - 1) / y) - 1;
+    }
+    return x / y;
+}
+
+fn rem_euclid(x: i32, y: i32) -> i32 {
+    return x - div_euclid(x, y) * y;
+}
diff --git a/src/client/render/tile/compute.wgsl b/src/client/render/tile/compute.wgsl
deleted file mode 100644
index 6556ee0..0000000
--- a/src/client/render/tile/compute.wgsl
+++ /dev/null
@@ -1,35 +0,0 @@
-
-@fragment
-fn fs_main(
-    in: VertexOutput,
-) -> @location(0) vec4<f32> {
-    let dec = i32(1) << 13;
-    let c = vec2<i32>(in.world_pos * f32(dec));
-    let cx = c.x;
-    let cy = c.y;
-    var x = 0;
-    var y = 0;
-    var i = 0u;
-    let thresh = 2 * dec;
-    let thresh2 = thresh * thresh;
-    let max = 50u;
-    loop {
-        let x2 = x * x;
-        let y2 = y * y;
-        if x2 + y2 > thresh2 || i >= max {
-            break;
-        }
-        y = (2 * x * y) / dec + c.y;
-        x = (x2 - y2) / dec + c.x;
-        i += 1u;
-    }
-    var color = vec3<f32>(0.0, 0.0, 0.0);
-    if i != max {
-        let pi = 3.1415;
-        let hue = f32(i) / 30.0;
-        color.r = cos(hue);
-        color.g = cos(hue - 2.0 * pi / 3.0);
-        color.b = cos(hue - 4.0 * pi / 3.0);
-    }
-    return vec4(color, 1.0);
-}
diff --git a/src/client/render/tile/layout.rs b/src/client/render/tile/layout.rs
deleted file mode 100644
index 35bd404..0000000
--- a/src/client/render/tile/layout.rs
+++ /dev/null
@@ -1,90 +0,0 @@
-use super::{util::Storage, View};
-
-pub struct TileLayout {
-    render_bind_layout: wgpu::BindGroupLayout,
-    render_pipeline_layout: wgpu::PipelineLayout,
-    format: wgpu::TextureFormat,
-    pub view: Storage,
-}
-
-impl TileLayout {
-    pub fn init(device: &wgpu::Device, config: &wgpu::SurfaceConfiguration) -> Self {
-        let view = Storage::init_with(
-            device,
-            "view",
-            View::default().bytes(),
-        );
-
-        let render_bind_layout =
-            device.create_bind_group_layout(&wgpu::BindGroupLayoutDescriptor {
-                entries: &[view.bind_group_layout_entry(0, true)],
-                label: Some("voxel render"),
-            });
-
-        let render_pipeline_layout =
-            device.create_pipeline_layout(&wgpu::PipelineLayoutDescriptor {
-                label: Some("Tile Pipeline Layout"),
-                bind_group_layouts: &[&render_bind_layout],
-                push_constant_ranges: &[],
-            });
-
-        Self {
-            view,
-            render_bind_layout,
-            render_pipeline_layout,
-            format: config.format,
-        }
-    }
-
-    pub fn render_bind_group(&self, device: &wgpu::Device) -> wgpu::BindGroup {
-        device.create_bind_group(&wgpu::BindGroupDescriptor {
-            layout: &self.render_bind_layout,
-            entries: &[self.view.bind_group_entry(0)],
-            label: Some("voxel render"),
-        })
-    }
-
-    pub fn render_pipeline(
-        &self,
-        device: &wgpu::Device,
-        shader: wgpu::ShaderModule,
-    ) -> wgpu::RenderPipeline {
-        device.create_render_pipeline(&wgpu::RenderPipelineDescriptor {
-            label: Some("Voxel Pipeline"),
-            layout: Some(&self.render_pipeline_layout),
-            vertex: wgpu::VertexState {
-                module: &shader,
-                entry_point: "vs_main",
-                buffers: &[],
-                compilation_options: wgpu::PipelineCompilationOptions::default(),
-            },
-            fragment: Some(wgpu::FragmentState {
-                module: &shader,
-                entry_point: "fs_main",
-                targets: &[Some(wgpu::ColorTargetState {
-                    format: self.format,
-                    blend: Some(wgpu::BlendState::REPLACE),
-                    write_mask: wgpu::ColorWrites::ALL,
-                })],
-                compilation_options: wgpu::PipelineCompilationOptions::default(),
-            }),
-            primitive: wgpu::PrimitiveState {
-                topology: wgpu::PrimitiveTopology::TriangleStrip,
-                strip_index_format: None,
-                front_face: wgpu::FrontFace::Ccw,
-                cull_mode: None,
-                polygon_mode: wgpu::PolygonMode::Fill,
-                unclipped_depth: false,
-                conservative: false,
-            },
-            depth_stencil: None,
-            multisample: wgpu::MultisampleState {
-                count: 1,
-                mask: !0,
-                alpha_to_coverage_enabled: true,
-            },
-            multiview: None,
-            cache: None,
-        })
-    }
-}
diff --git a/src/client/render/tile/mod.rs b/src/client/render/tile/mod.rs
deleted file mode 100644
index 2e8ba0a..0000000
--- a/src/client/render/tile/mod.rs
+++ /dev/null
@@ -1,61 +0,0 @@
-use std::ops::{Deref, DerefMut};
-
-use wgpu::include_wgsl;
-
-mod data;
-mod layout;
-
-use super::*;
-pub use data::*;
-use layout::*;
-
-pub struct TilePipeline {
-    layout: TileLayout,
-    render_pipeline: wgpu::RenderPipeline,
-    render_bind_group: wgpu::BindGroup,
-}
-
-const RENDER_SHADER: wgpu::ShaderModuleDescriptor<'_> = include_wgsl!("render.wgsl");
-
-impl TilePipeline {
-    pub fn init(device: &wgpu::Device, config: &wgpu::SurfaceConfiguration) -> Self {
-        let layout = TileLayout::init(device, config);
-        let render_shader = device.create_shader_module(RENDER_SHADER);
-        Self {
-            render_pipeline: layout.render_pipeline(device, render_shader),
-            render_bind_group: layout.render_bind_group(device),
-            layout,
-        }
-    }
-
-    pub fn update(
-        &mut self,
-        device: &wgpu::Device,
-        encoder: &mut wgpu::CommandEncoder,
-        belt: &mut wgpu::util::StagingBelt,
-        camera: &Camera,
-        size: &Vector2<u32>,
-    ) {
-        self.view.update(device, encoder, belt, View::from_camera_size(camera, size).bytes());
-    }
-
-    pub fn draw<'a>(&'a self, render_pass: &mut wgpu::RenderPass<'a>) {
-        render_pass.set_pipeline(&self.render_pipeline);
-        render_pass.set_bind_group(0, &self.render_bind_group, &[]);
-        render_pass.draw(0..4, 0..1);
-    }
-}
-
-impl Deref for TilePipeline {
-    type Target = TileLayout;
-
-    fn deref(&self) -> &Self::Target {
-        &self.layout
-    }
-}
-
-impl DerefMut for TilePipeline {
-    fn deref_mut(&mut self) -> &mut Self::Target {
-        &mut self.layout
-    }
-}
diff --git a/src/client/render/util/array.rs b/src/client/render/util/array.rs
index 74a534d..cca5a60 100644
--- a/src/client/render/util/array.rs
+++ b/src/client/render/util/array.rs
@@ -6,7 +6,7 @@ pub struct ArrayBuffer<T: bytemuck::Pod> {
     buffer: wgpu::Buffer,
     label: String,
     usage: BufferUsages,
-    updates: Vec<ArrBufUpdate<T>>,
+    update: Option<Vec<T>>,
 }
 
 impl<T: bytemuck::Pod> ArrayBuffer<T> {
@@ -36,34 +36,24 @@ impl<T: bytemuck::Pod> ArrayBuffer<T> {
         if self.len == 0 {
             return resized;
         }
-        for update in &self.updates {
+        if let Some(update) = self.update.take() {
             let mut view = belt.write_buffer(
                 encoder,
                 &self.buffer,
-                (update.offset * std::mem::size_of::<T>()) as BufferAddress,
+                0,
                 unsafe {
-                    std::num::NonZeroU64::new_unchecked(
-                        std::mem::size_of_val(&update.data[..]) as u64
-                    )
+                    std::num::NonZeroU64::new_unchecked(std::mem::size_of_val(&update[..]) as u64)
                 },
                 device,
             );
-            view.copy_from_slice(bytemuck::cast_slice(&update.data));
+            view.copy_from_slice(bytemuck::cast_slice(&update));
         }
-        self.updates.clear();
         resized
     }
 
-    pub fn add(&mut self, data: Vec<T>) -> usize {
-        let data_len = data.len();
-        let pos = self.new_len;
-        self.updates.push(ArrBufUpdate { offset: pos, data });
-        self.new_len += data_len;
-        pos
-    }
-
-    pub fn set(&mut self, offset: usize, data: Vec<T>) {
-        self.updates.push(ArrBufUpdate { offset, data });
+    pub fn set(&mut self, data: Vec<T>) {
+        self.new_len = data.len();
+        self.update = Some(data);
     }
 
     pub fn init(device: &wgpu::Device, label: &str, usage: BufferUsages) -> Self {
@@ -82,7 +72,7 @@ impl<T: bytemuck::Pod> ArrayBuffer<T> {
                 Self::init_buf_with(device, label, usage, data)
             },
             label: label.to_string(),
-            updates: Vec::new(),
+            update: None,
             usage,
         }
     }
@@ -117,8 +107,8 @@ impl<T: bytemuck::Pod> ArrayBuffer<T> {
     pub fn bind_group_layout_entry(
         &self,
         binding: u32,
-        visibility: wgpu::ShaderStages,
         ty: wgpu::BufferBindingType,
+        visibility: wgpu::ShaderStages,
     ) -> wgpu::BindGroupLayoutEntry {
         wgpu::BindGroupLayoutEntry {
             binding,
diff --git a/src/client/render/util/storage.rs b/src/client/render/util/storage.rs
index d36fdfc..daa3956 100644
--- a/src/client/render/util/storage.rs
+++ b/src/client/render/util/storage.rs
@@ -4,56 +4,68 @@ use wgpu::util::DeviceExt;
 
 pub struct Storage {
     buffer: wgpu::Buffer,
+    old_len: usize,
 }
 
 impl Storage {
     pub fn init<T: Default + bytemuck::Pod>(device: &wgpu::Device, name: &str) -> Self {
+        let def = [T::default()];
+        let default = bytemuck::cast_slice(&def);
         Self {
             buffer: device.create_buffer_init(&wgpu::util::BufferInitDescriptor {
                 label: Some(&(name.to_owned() + " Uniform Buf")),
-                contents: bytemuck::cast_slice(&[T::default()]),
+                contents: default,
                 usage: wgpu::BufferUsages::STORAGE | wgpu::BufferUsages::COPY_DST,
             }),
+            old_len: default.len(),
         }
     }
     pub fn init_with(device: &wgpu::Device, name: &str, data: &[u8]) -> Self {
         Self {
-            buffer: device.create_buffer_init(&wgpu::util::BufferInitDescriptor {
-                label: Some(&(name.to_owned() + " Uniform Buf")),
-                contents: data,
-                usage: wgpu::BufferUsages::STORAGE | wgpu::BufferUsages::COPY_DST,
-            }),
+            buffer: Self::init_buf(device, name, data),
+            old_len: data.len(),
         }
     }
+    pub fn init_buf(device: &wgpu::Device, name: &str, data: &[u8]) -> wgpu::Buffer {
+        device.create_buffer_init(&wgpu::util::BufferInitDescriptor {
+            label: Some(&(name.to_owned() + " Uniform Buf")),
+            contents: data,
+            usage: wgpu::BufferUsages::STORAGE | wgpu::BufferUsages::COPY_DST,
+        })
+    }
     pub fn update(
         &mut self,
         device: &wgpu::Device,
         encoder: &mut wgpu::CommandEncoder,
         belt: &mut wgpu::util::StagingBelt,
         data: &[u8],
-    ) {
-        let mut view = belt.write_buffer(
-            encoder,
-            &self.buffer,
-            0,
-            unsafe {
-                std::num::NonZeroU64::new_unchecked(std::mem::size_of_val(data) as u64)
-            },
-            device,
-        );
-        view.copy_from_slice(data);
+    ) -> bool {
+        if data.len() != self.old_len {
+            self.buffer = Self::init_buf(device, "too lazy", data);
+            self.old_len = data.len();
+            true
+        } else {
+            let mut view = belt.write_buffer(
+                encoder,
+                &self.buffer,
+                0,
+                unsafe { std::num::NonZeroU64::new_unchecked(std::mem::size_of_val(data) as u64) },
+                device,
+            );
+            view.copy_from_slice(data);
+            false
+        }
     }
 
     pub fn bind_group_layout_entry(
         &self,
         binding: u32,
         read_only: bool,
+        visibility: wgpu::ShaderStages,
     ) -> wgpu::BindGroupLayoutEntry {
         wgpu::BindGroupLayoutEntry {
             binding,
-            visibility: wgpu::ShaderStages::VERTEX
-                | wgpu::ShaderStages::FRAGMENT
-                | wgpu::ShaderStages::COMPUTE,
+            visibility,
             ty: wgpu::BindingType::Buffer {
                 ty: wgpu::BufferBindingType::Storage { read_only },
                 has_dynamic_offset: false,
diff --git a/src/client/render/util/texture.rs b/src/client/render/util/texture.rs
index 49bd121..9b8a740 100644
--- a/src/client/render/util/texture.rs
+++ b/src/client/render/util/texture.rs
@@ -89,3 +89,39 @@ impl Texture {
         self.texture_desc.format
     }
 }
+
+pub struct ResizableTexture {
+    texture_desc: wgpu::TextureDescriptor<'static>,
+    view_desc: wgpu::TextureViewDescriptor<'static>,
+    pub texture: wgpu::Texture,
+    pub view: wgpu::TextureView,
+}
+
+impl ResizableTexture {
+    pub fn new(
+        device: &wgpu::Device,
+        texture_desc: wgpu::TextureDescriptor<'static>,
+        view_desc: wgpu::TextureViewDescriptor<'static>,
+    ) -> Self {
+        let texture = device.create_texture(&texture_desc);
+        let view = texture.create_view(&view_desc);
+        Self {
+            texture,
+            view,
+            texture_desc,
+            view_desc,
+        }
+    }
+    pub fn resize_layers(&mut self, device: &wgpu::Device, layers: u32) {
+        self.texture_desc.size.depth_or_array_layers = layers;
+        self.texture = device.create_texture(&self.texture_desc);
+        self.view = self.texture.create_view(&self.view_desc)
+    }
+
+    pub fn view_entry(&self, binding: u32) -> wgpu::BindGroupEntry {
+        wgpu::BindGroupEntry {
+            binding,
+            resource: wgpu::BindingResource::TextureView(&self.view),
+        }
+    }
+}
diff --git a/src/client/render/view.rs b/src/client/render/view.rs
new file mode 100644
index 0000000..b746ad1
--- /dev/null
+++ b/src/client/render/view.rs
@@ -0,0 +1,66 @@
+use std::collections::HashSet;
+
+use nalgebra::Vector2;
+
+use crate::{client::camera::Camera, util::FixedDec};
+
+use super::{output::WindowView, CHUNK_POW};
+
+#[derive(Default)]
+pub struct ChunkView {
+    pub render: WindowView,
+    pub chunk_queue: HashSet<Vector2<FixedDec>>,
+    pub visible_chunks: HashSet<Vector2<FixedDec>>,
+}
+
+impl ChunkView {
+    pub fn new() -> Self {
+        Self::default()
+    }
+
+    pub fn update(&mut self, camera: &Camera, size: &Vector2<u32>) {
+        let render = WindowView::from_camera_size(camera, size);
+
+        if self.render == render {
+            return;
+        }
+        self.render = render;
+
+        let corner_offset = ((size / 2).cast() * camera.zoom.rel_zoom())
+            .map(|x| FixedDec::from(x) >> camera.zoom.level());
+        let bot_left = &camera.pos - &corner_offset;
+        let top_right = &camera.pos + &corner_offset;
+        let mult = FixedDec::one() >> (CHUNK_POW as i32 - camera.zoom.level());
+        let blc = bot_left
+            .component_mul(&Vector2::from_element(mult.clone()))
+            .map(FixedDec::floor);
+        let trc = top_right
+            .component_mul(&Vector2::from_element(mult))
+            .map(FixedDec::floor);
+
+        let mut visible = HashSet::new();
+        let mut x = blc.x.clone();
+        while x <= trc.x {
+            let mut y = blc.y.clone();
+            while y <= trc.y {
+                visible.insert(Vector2::new(x.clone(), y.clone()));
+                y += FixedDec::one();
+            }
+            x += FixedDec::one();
+        }
+
+        let new = visible
+            .difference(&self.visible_chunks)
+            .cloned()
+            .collect::<Vec<_>>();
+        let old = self
+            .visible_chunks
+            .difference(&visible)
+            .cloned()
+            .collect::<Vec<_>>();
+        self.chunk_queue.retain(|p| !old.contains(p));
+        self.chunk_queue.extend(new);
+        self.visible_chunks = visible;
+    }
+}
+
diff --git a/src/util/fixed/conversion.rs b/src/util/fixed/conversion.rs
index f33cfbf..a6642eb 100644
--- a/src/util/fixed/conversion.rs
+++ b/src/util/fixed/conversion.rs
@@ -8,6 +8,9 @@ const FRAC_MASK: u32 = FRAC_BIT - 1;
 
 impl From<f32> for FixedDec {
     fn from(value: f32) -> Self {
+        if value.is_zero() {
+            return Self::zero();
+        }
         let raw = value.to_bits() & INV_SIGN_MASK;
         let mut exp = (raw >> 23) as i32 - 127;
         let mut frac = raw & FRAC_MASK;
diff --git a/src/util/fixed/mod.rs b/src/util/fixed/mod.rs
index eba94f8..707669d 100644
--- a/src/util/fixed/mod.rs
+++ b/src/util/fixed/mod.rs
@@ -17,7 +17,7 @@ const NEG: bool = true;
 // Might want to try to make it abstract over the direction
 // of growth, or use a VecDeque, but doesn't seem worth for
 // now
-#[derive(Debug, Clone, PartialEq)]
+#[derive(Debug, Clone, PartialEq, Eq, Hash)]
 pub struct FixedDec {
     sign: bool,
     dec: i32,
@@ -25,6 +25,14 @@ pub struct FixedDec {
 }
 
 impl FixedDec {
+    pub fn one() -> Self {
+        Self {
+            sign: POS,
+            dec: 1,
+            parts: vec![1],
+        }
+    }
+
     pub fn from_parts(sign: bool, dec: i32, parts: Vec<u32>) -> Self {
         Self { sign, dec, parts }
     }
@@ -62,18 +70,26 @@ impl FixedDec {
         } else {
             self.dec -= rem_beg as i32;
         }
+        if self.parts.is_empty() {
+            self.sign = POS;
+        }
     }
 
     pub fn set_whole_len(&mut self, len: i32) {
         let diff = len - self.dec;
-        self.parts
-            .splice(0..usize::try_from(-diff).unwrap_or(0), (0..diff).map(|_| 0));
+        let remove = 0..usize::try_from(-diff).unwrap_or(0).min(self.parts.len());
+        self.parts.splice(remove, (0..diff).map(|_| 0));
         self.dec += diff;
         if self.parts.is_empty() {
             self.dec = 0;
         }
     }
 
+    pub fn with_whole_len(mut self, len: i32) -> Self {
+        self.set_whole_len(len);
+        self
+    }
+
     pub fn set_dec_len(&mut self, len: i32) {
         let len = usize::try_from(len + self.dec).unwrap_or(0);
         self.parts.resize(len, 0);
@@ -82,6 +98,31 @@ impl FixedDec {
         }
     }
 
+    pub fn with_dec_len(mut self, len: i32) -> Self {
+        self.set_dec_len(len);
+        self
+    }
+
+    pub fn with_lens(self, whole_len: i32, dec_len: i32) -> Self {
+        self.with_whole_len(whole_len).with_dec_len(dec_len)
+    }
+
+    pub fn split_whole_dec(&self) -> (FixedDec, FixedDec) {
+        let take_skip = usize::try_from(self.dec).unwrap_or(0);
+        let whole = FixedDec {
+            sign: self.sign,
+            dec: self.dec,
+            parts: self.parts.iter().take(take_skip).copied().collect(),
+        };
+        let dec = if self.dec >= 0 { 0 } else { self.dec };
+        let dec = FixedDec {
+            sign: self.sign,
+            dec,
+            parts: self.parts.iter().skip(take_skip).copied().collect(),
+        };
+        (whole, dec)
+    }
+
     pub fn set_precision(&mut self, prec: usize) {
         self.parts.resize(prec, 0);
         if self.parts.is_empty() {
@@ -94,6 +135,10 @@ impl FixedDec {
         bytes.extend(self.dec.to_le_bytes());
         bytes.extend(self.parts.iter().flat_map(|p| p.to_le_bytes()));
     }
+
+    pub fn parts(&self) -> &[u32] {
+        &self.parts
+    }
 }
 
 impl Display for FixedDec {
diff --git a/src/util/fixed/op.rs b/src/util/fixed/op.rs
index 3593f9a..1a38437 100644
--- a/src/util/fixed/op.rs
+++ b/src/util/fixed/op.rs
@@ -1,6 +1,9 @@
 use super::*;
 
-use std::ops::{Add, AddAssign, Mul, Neg, Shr, Sub, SubAssign};
+use std::{
+    cmp::Ordering,
+    ops::{Add, AddAssign, Mul, MulAssign, Neg, Shl, Shr, Sub, SubAssign},
+};
 
 impl Zero for FixedDec {
     fn zero() -> Self {
@@ -16,8 +19,29 @@ impl Zero for FixedDec {
     }
 }
 
+impl Shl<i32> for FixedDec {
+    type Output = FixedDec;
+    fn shl(self, rhs: i32) -> Self::Output {
+        &self << rhs
+    }
+}
+
+impl Shl<i32> for &FixedDec {
+    type Output = FixedDec;
+    fn shl(self, rhs: i32) -> Self::Output {
+        self >> -rhs
+    }
+}
+
 impl Shr<i32> for FixedDec {
-    type Output = Self;
+    type Output = FixedDec;
+    fn shr(self, rhs: i32) -> Self::Output {
+        &self >> rhs
+    }
+}
+
+impl Shr<i32> for &FixedDec {
+    type Output = FixedDec;
 
     fn shr(self, rhs: i32) -> Self::Output {
         let mut parts = Vec::with_capacity(self.parts.len());
@@ -26,14 +50,14 @@ impl Shr<i32> for FixedDec {
         let mask = (1 << sr) - 1;
         let dec = self.dec - rhs.div_floor(32);
         let mut rem = 0;
-        for part in self.parts {
+        for part in &self.parts {
             parts.push((part >> sr) ^ rem);
             rem = (part & mask).unbounded_shl(sl);
         }
         if rem != 0 {
             parts.push(rem);
         }
-        Self {
+        Self::Output {
             dec,
             parts,
             sign: self.sign,
@@ -124,6 +148,7 @@ fn add(dest: &mut FixedDec, src: &impl Fn(usize) -> u32, rhs: &FixedDec) {
             }
         }
     }
+    dest.trim()
 }
 
 fn new_dec(x: &FixedDec, y: &FixedDec) -> (i32, usize) {
@@ -157,6 +182,12 @@ impl SubAssign<&FixedDec> for FixedDec {
     }
 }
 
+impl SubAssign<FixedDec> for FixedDec {
+    fn sub_assign(&mut self, rhs: FixedDec) {
+        *self += &-rhs;
+    }
+}
+
 impl Neg for &FixedDec {
     type Output = FixedDec;
 
@@ -230,6 +261,18 @@ impl Mul<FixedDec> for &FixedDec {
     }
 }
 
+impl MulAssign<&FixedDec> for FixedDec {
+    fn mul_assign(&mut self, rhs: &FixedDec) {
+        *self = &*self * rhs
+    }
+}
+
+impl MulAssign<FixedDec> for FixedDec {
+    fn mul_assign(&mut self, rhs: FixedDec) {
+        *self = &*self * rhs
+    }
+}
+
 fn mul_lmsb(x: u32, y: u32) -> (u32, u32) {
     let lsb = x.wrapping_mul(y);
     let a = x & 0xffff;
@@ -242,3 +285,65 @@ fn mul_lmsb(x: u32, y: u32) -> (u32, u32) {
     let msb = ((ad + bc) >> 16) + ((carry as u32) << 16) + b * d;
     (lsb, msb)
 }
+
+impl PartialOrd for FixedDec {
+    fn partial_cmp(&self, other: &Self) -> Option<Ordering> {
+        Some(self.cmp(other))
+    }
+}
+
+impl Ord for FixedDec {
+    fn cmp(&self, other: &Self) -> std::cmp::Ordering {
+        match (self.sign, other.sign) {
+            (POS, NEG) => return Ordering::Greater,
+            (NEG, POS) => return Ordering::Less,
+            _ => (),
+        }
+        match (self.dec.cmp(&other.dec), self.sign) {
+            (Ordering::Less, POS) => return Ordering::Less,
+            (Ordering::Less, NEG) => return Ordering::Greater,
+            (Ordering::Greater, POS) => return Ordering::Greater,
+            (Ordering::Greater, NEG) => return Ordering::Less,
+            (Ordering::Equal, _) => (),
+        }
+        match (self.parts.first().unwrap_or(&0).cmp(other.parts.first().unwrap_or(&0)), self.sign) {
+            (Ordering::Less, POS) => Ordering::Less,
+            (Ordering::Less, NEG) => Ordering::Greater,
+            (Ordering::Greater, POS) => Ordering::Greater,
+            (Ordering::Greater, NEG) => Ordering::Less,
+            (Ordering::Equal, _) => Ordering::Equal,
+        }
+    }
+}
+
+impl FixedDec {
+    pub fn floor(mut self) -> Self {
+        if self.sign == NEG {
+            let diff = self.parts.len() as i32 - self.dec;
+            if diff > 0
+                && self.parts[self.dec.max(0) as usize..]
+                    .iter()
+                    .any(|p| *p != 0)
+            {
+                self -= Self::one();
+            }
+        }
+        self.parts.truncate(self.dec.max(0) as usize);
+        self
+    }
+    pub fn ceil(mut self) -> Self {
+        if self.sign == NEG {
+            let diff = self.parts.len() as i32 - self.dec;
+            if diff > 0
+                && self.parts[self.dec.max(0) as usize..]
+                    .iter()
+                    .any(|p| *p != 0)
+            {
+                self -= Self::one();
+            }
+        }
+        self.parts.truncate(self.dec.max(0) as usize);
+        self += Self::one();
+        self
+    }
+}
diff --git a/src/util/fixed/test.rs b/src/util/fixed/test.rs
index 9ad90da..d987e4e 100644
--- a/src/util/fixed/test.rs
+++ b/src/util/fixed/test.rs
@@ -15,7 +15,7 @@ macro_rules! assert_eq_bits {
     ($left:expr, $left_bits:expr, $right:expr, $right_bits:expr, $dec:expr $(,)?) => {
         assert!(
             ($left - $right).abs() < EPSILON,
-            "\n  left: {:032b} = {:?}\n right: {:032b} = {:?}\n  from: {:?}",
+            "\n  expect: {:032b} = {:?}\n found: {:032b} = {:?}\n  from: {:?}",
             $left_bits,
             $left,
             $right_bits,
@@ -26,7 +26,7 @@ macro_rules! assert_eq_bits {
     ($left:expr, $left_bits:expr, $right:expr, $right_bits:expr, $dec:expr, $arg:tt, $($args:tt)+) => {
         assert!(
             ($left - $right).abs() < EPSILON,
-            concat!("\n  expr: ", $arg, "\n  left: {:032b} = {:?}\n right: {:032b} = {:?}\n  from: {:?}"),
+            concat!("\n  expr: ", $arg, "\nexpect: {:032b} = {:?}\n found: {:032b} = {:?}\n  from: {:?}"),
             $($args)+,
             $left_bits,
             $left,
@@ -71,6 +71,7 @@ fn add_sub() {
         let dec = FixedDec::from(x) + FixedDec::from(y);
         assert_eq_f32!(a, f32::from(&dec), dec, "{} + {}", x, y);
     }
+    test(-0.0, 1.0);
     test(0.25, 0.75);
     test(1.25, 0.125);
     test(1.25, -0.125);
@@ -130,3 +131,23 @@ fn shr() {
     test(1, 33);
     test(1, -33);
 }
+
+#[test]
+fn ord() {
+    fn test(x: f32, y: f32) {
+        let l = FixedDec::from(x);
+        let r = FixedDec::from(y);
+        assert_eq!(x > y, l > r, "{:?} > {:?}", x, y);
+        assert_eq!(x < y, l < r, "{:?} > {:?}", x, y);
+        assert_eq!(x == y, l == r, "{:?} > {:?}", x, y);
+    }
+    test(1.0, 3.0);
+    test(1.0, -3.0);
+    test(-1.0, 3.0);
+    test(-1.0, -33.0);
+    test(-1.5, -33.0);
+    test(-1.5, 66.0);
+    test(-1.5, -1.5);
+    test(0.0, 0.0);
+    test(1.5000, -33.0);
+}