support for emscripten build

Cargo.toml

@@ -22,18 +22,34 @@ default = []
 #metal-auto-capture = ["gfx-backend-metal/auto-capture"]
 #vulkan-portability = ["wgc/gfx-backend-vulkan"]
 
-[dependencies.wgc]
+[target.'cfg(not(target_os = "emscripten"))'.dependencies.wgc]
 package = "wgpu-core"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "ec1d022"
+rev = "8b66507"
 # path = "../wgpu/wgpu-core"
 #version = "0.11"
 features = ["raw-window-handle", "trace"]
 
+[target.'cfg(target_os = "emscripten")'.dependencies.wgc]
+package = "wgpu-core"
+git = "https://github.com/gfx-rs/wgpu"
+rev = "8b66507"
+# path = "../wgpu/wgpu-core"
+#version = "0.11"
+features = ["raw-window-handle", "id32"]
+
+[target.'cfg(target_os = "emscripten")'.dependencies.hal]
+package = "wgpu-hal"
+git = "https://github.com/gfx-rs/wgpu"
+rev = "8b66507"
+# path = "../wgpu/wgpu-core"
+#version = "0.11"
+features = ["emscripten"]
+
 [dependencies.wgt]
 package = "wgpu-types"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "ec1d022"
+rev = "8b66507"
 # path = "../wgpu/wgpu-types"
 #version = "0.11"
 
@@ -46,7 +62,7 @@ thiserror = "1"
 
 [dependencies.naga]
 git = "https://github.com/gfx-rs/naga"
-rev = "8ffd6ba"
+rev = "8e2e39e"
 features = ["spv-in"]
 
 [build-dependencies]

Makefile

@@ -101,6 +101,9 @@ lib-native: Cargo.lock Cargo.toml Makefile $(WILDCARD_SOURCE)
 lib-native-release: Cargo.lock Cargo.toml Makefile $(WILDCARD_SOURCE)
 	cargo build --release $(EXTRA_BUILD_ARGS)
 
+lib-native-emscripten: Cargo.lock Cargo.toml Makefile $(WILDCARD_SOURCE)
+	EMCC_CFLAGS="-Oz -g -s ERROR_ON_UNDEFINED_SYMBOLS=0 --no-entry -s ASSERTIONS=0" cargo build --target wasm32-unknown-emscripten --release $(EXTRA_BUILD_ARGS)
+
 example-compute: lib-native examples/compute/main.c
 	cd examples/compute && $(CREATE_BUILD_DIR) && cd build && cmake -DCMAKE_BUILD_TYPE=Debug .. $(GENERATOR_PLATFORM) && cmake --build .
 
@@ -113,6 +116,9 @@ example-triangle: lib-native examples/triangle/main.c
 run-example-triangle: example-triangle
 	cd examples/triangle && "$(OUTPUT_DIR)/triangle"
 
+example-triangle-emscripten: lib-native-emscripten examples/triangle/main.c
+	cd examples/triangle && $(CREATE_BUILD_DIR) && cd build && emcmake cmake -DCMAKE_BUILD_TYPE=Debug .. && cmake --build .
+
 build-helper:
 	cargo build -p helper
 

examples/compute/shader.wgsl

@@ -1,8 +1,9 @@
 struct PrimeIndices {
-    data: [[stride(4)]] array<u32>;
+    data: @stride(4) array<u32>;
 }; // this is used as both input and output for convenience
 
-[[group(0), binding(0)]]
+@group(0)
+@binding(0)
 var<storage, read_write> v_indices: PrimeIndices;
 
 // The Collatz Conjecture states that for any integer n:
@@ -34,7 +35,8 @@ fn collatz_iterations(n_base: u32) -> u32{
     return i;
 }
 
-[[stage(compute), workgroup_size(1)]]
-fn main([[builtin(global_invocation_id)]] global_id: vec3<u32>) {
+@stage(compute)
+@workgroup_size(1)
+fn main(@builtin(global_invocation_id) global_id: vec3<u32>) {
     v_indices.data[global_id.x] = collatz_iterations(v_indices.data[global_id.x]);
 }

examples/triangle/CMakeLists.txt

@@ -6,7 +6,9 @@ set(TARGET_NAME triangle)
 
 add_executable(triangle main.c ../framework.c)
 
-if(MSVC)
+if(EMSCRIPTEN)
+    add_definitions(-DWGPU_TARGET=WGPU_TARGET_LINUX_EMSCRIPTEN)
+elseif(MSVC)
     add_definitions(-DWGPU_TARGET=WGPU_TARGET_WINDOWS)
     target_compile_options(${TARGET_NAME} PRIVATE /W4)
     set(OS_LIBRARIES "userenv" "ws2_32" "Dwmapi" "dbghelp" "d3dcompiler" "D3D12" "D3D11" "DXGI" "setupapi" "Bcrypt")
@@ -23,15 +25,27 @@ endif(USE_WAYLAND)
     target_compile_options(${TARGET_NAME} PRIVATE -Wall -Wextra -pedantic)
 endif(MSVC)
 
-find_package(glfw3 3.3 REQUIRED
-    HINTS "$ENV{GLFW3_INSTALL_DIR}"
-)
+if(EMSCRIPTEN)
+    set(CMAKE_C_FLAGS "${CMAKE_C_FLAGS} \
+        -Oz -g -s USE_GLFW=3")
+    set(CMAKE_EXE_LINKER_FLAGS "${CMAKE_EXE_LINKER_FLAGS} \
+        -Oz -g -s FULL_ES2=1 \
+        -s ERROR_ON_UNDEFINED_SYMBOLS=0 \
+        -s MIN_WEBGL_VERSION=2 \
+        --preload-file \"${CMAKE_CURRENT_SOURCE_DIR}/shader.wgsl@shader.wgsl\"")
+    set(WGPU_LIBRARY "wgpu_native")
+    set_target_properties(${TARGET_NAME} PROPERTIES SUFFIX .html)
+    target_link_directories(${TARGET_NAME} PUBLIC "${CMAKE_CURRENT_SOURCE_DIR}/../../target/wasm32-unknown-emscripten/release")
+else(EMSCRIPTEN)
+    find_package(glfw3 3.3 REQUIRED
+        HINTS "$ENV{GLFW3_INSTALL_DIR}"
+    )
+    find_library(WGPU_LIBRARY wgpu_native
+        HINTS "${CMAKE_CURRENT_SOURCE_DIR}/../../target/debug"
+    )
+    target_include_directories(${TARGET_NAME} PUBLIC $ENV{GLFW3_INCLUDE_DIR})
+endif(EMSCRIPTEN)
 
-find_library(WGPU_LIBRARY wgpu_native
-    HINTS "${CMAKE_CURRENT_SOURCE_DIR}/../../target/debug"
-)
-
-target_include_directories(${TARGET_NAME} PUBLIC $ENV{GLFW3_INCLUDE_DIR})
 target_include_directories(${TARGET_NAME} PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/../../ffi)
 target_include_directories(${TARGET_NAME} PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/../)
 

examples/triangle/main.c

@@ -25,7 +25,9 @@
 #elif WGPU_TARGET == WGPU_TARGET_WINDOWS
 #define GLFW_EXPOSE_NATIVE_WIN32
 #endif
+#if WGPU_TARGET != WGPU_TARGET_LINUX_EMSCRIPTEN
 #include <GLFW/glfw3native.h>
+#endif
 
 static void handle_device_lost(WGPUDeviceLostReason reason, char const * message, void * userdata)
 {
@@ -45,7 +47,11 @@ int main() {
     return 1;
   }
 
+#if WGPU_TARGET == WGPU_TARGET_LINUX_EMSCRIPTEN
+  glfwWindowHint(GLFW_CLIENT_API, GLFW_OPENGL_ES_API);
+#else
   glfwWindowHint(GLFW_CLIENT_API, GLFW_NO_API);
+#endif
   GLFWwindow *window = glfwCreateWindow(640, 480, "wgpu with glfw", NULL, NULL);
 
   if (!window) {
@@ -54,8 +60,16 @@ int main() {
   }
 
   WGPUSurface surface;
-
-#if WGPU_TARGET == WGPU_TARGET_MACOS
+#if WGPU_TARGET == WGPU_TARGET_LINUX_EMSCRIPTEN
+  {
+    surface = wgpuInstanceCreateSurface(
+        NULL,
+        &(WGPUSurfaceDescriptor){
+            .label = NULL,
+            .nextInChain = NULL,
+        });
+  }
+#elif WGPU_TARGET == WGPU_TARGET_MACOS
   {
     id metal_layer = NULL;
     NSWindow *ns_window = glfwGetCocoaWindow(window);
@@ -332,6 +346,11 @@ int main() {
     wgpuSwapChainPresent(swapChain);
 
     glfwPollEvents();
+
+#if WGPU_TARGET == WGPU_TARGET_LINUX_EMSCRIPTEN
+    // you must use emscripten_set_main_loop or asyncify
+    return 0;
+#endif
   }
 
   glfwDestroyWindow(window);

examples/triangle/shader.wgsl

@@ -1,11 +1,11 @@
-[[stage(vertex)]]
-fn vs_main([[builtin(vertex_index)]] in_vertex_index: u32) -> [[builtin(position)]] vec4<f32> {
+@stage(vertex)
+fn vs_main(@builtin(vertex_index) in_vertex_index: u32) -> @builtin(position) vec4<f32> {
     let x = f32(i32(in_vertex_index) - 1);
     let y = f32(i32(in_vertex_index & 1u) * 2 - 1);
     return vec4<f32>(x, y, 0.0, 1.0);
 }
 
-[[stage(fragment)]]
-fn fs_main() -> [[location(0)]] vec4<f32> {
+@stage(fragment)
+fn fs_main() -> @location(0) vec4<f32> {
     return vec4<f32>(1.0, 0.0, 0.0, 1.0);
 }

src/conv.rs

@@ -227,7 +227,12 @@ pub fn map_device_descriptor<'a>(
 }
 
 pub fn map_limits(limits: native::WGPULimits) -> wgt::Limits {
+    #[cfg(target_os = "emscripten")]
+    let mut wgt_limits = wgt::Limits::downlevel_webgl2_defaults();
+
+    #[cfg(not(target_os = "emscripten"))]
     let mut wgt_limits = wgt::Limits::default();
+
     if limits.maxTextureDimension1D != 0 {
         wgt_limits.max_texture_dimension_1d = limits.maxTextureDimension1D;
     }
@@ -272,46 +277,42 @@ pub fn map_limits(limits: native::WGPULimits) -> wgt::Limits {
     if limits.maxStorageBufferBindingSize != 0 {
         wgt_limits.max_storage_buffer_binding_size = limits.maxStorageBufferBindingSize as u32;
     }
-    /* not yet available in wgpu-core
     if limits.minUniformBufferOffsetAlignment != 0 {
-        wgt_limits.yyyy = limits.minUniformBufferOffsetAlignment;
+        wgt_limits.min_uniform_buffer_offset_alignment = limits.minUniformBufferOffsetAlignment;
     }
     if limits.minStorageBufferOffsetAlignment != 0 {
-        wgt_limits.yyyy = limits.minStorageBufferOffsetAlignment;
+        wgt_limits.min_storage_buffer_offset_alignment = limits.minStorageBufferOffsetAlignment;
     }
-    */
     if limits.maxVertexBuffers != 0 {
         wgt_limits.max_vertex_buffers = limits.maxVertexBuffers;
     }
     if limits.maxVertexAttributes != 0 {
         wgt_limits.max_vertex_attributes = limits.maxVertexAttributes;
     }
-    /* not yet available in wgpu-core
     if limits.maxVertexBufferArrayStride != 0 {
-        wgt_limits.yyyy = limits.maxVertexBufferArrayStride;
+        wgt_limits.max_vertex_buffer_array_stride = limits.maxVertexBufferArrayStride;
     }
     if limits.maxInterStageShaderComponents != 0 {
-        wgt_limits.yyyy = limits.maxInterStageShaderComponents;
+        wgt_limits.max_inter_stage_shader_components = limits.maxInterStageShaderComponents;
     }
     if limits.maxComputeWorkgroupStorageSize != 0 {
-        wgt_limits.yyyy = limits.maxComputeWorkgroupStorageSize;
+        wgt_limits.max_compute_workgroup_storage_size = limits.maxComputeWorkgroupStorageSize;
     }
     if limits.maxComputeInvocationsPerWorkgroup != 0 {
-        wgt_limits.yyyy = limits.maxComputeInvocationsPerWorkgroup;
+        wgt_limits.max_compute_invocations_per_workgroup = limits.maxComputeInvocationsPerWorkgroup;
     }
     if limits.maxComputeWorkgroupSizeX != 0 {
-        wgt_limits.yyyy = limits.maxComputeWorkgroupSizeX;
+        wgt_limits.max_compute_workgroup_size_x = limits.maxComputeWorkgroupSizeX;
     }
     if limits.maxComputeWorkgroupSizeY != 0 {
-        wgt_limits.yyyy = limits.maxComputeWorkgroupSizeY;
+        wgt_limits.max_compute_workgroup_size_y = limits.maxComputeWorkgroupSizeY;
     }
     if limits.maxComputeWorkgroupSizeZ != 0 {
-        wgt_limits.yyyy = limits.maxComputeWorkgroupSizeZ;
+        wgt_limits.max_compute_workgroup_size_z = limits.maxComputeWorkgroupSizeZ;
     }
     if limits.maxComputeWorkgroupsPerDimension != 0 {
-        wgt_limits.yyyy = limits.maxComputeWorkgroupsPerDimension;
+        wgt_limits.max_compute_workgroups_per_dimension = limits.maxComputeWorkgroupsPerDimension;
     }
-    */
     return wgt_limits;
 }