Make misc fixes (#41)

* Update ze_loader.lib to the newest version
* Export _ptsz/_ptds for which we have a legacy stream implementations
* Stop producing build logs if we are not looking at them anyway

2 files changed, 80 insertions, 2 deletions

diff --git a/zluda/src/cuda.rs b/zluda/src/cuda.rs
index 469f8f3..1eb08d5 100644
--- a/zluda/src/cuda.rs
+++ b/zluda/src/cuda.rs
@@ -2454,6 +2454,7 @@ pub extern "C" fn cuModuleLoadData(
     r#impl::module::load_data(module.decuda(), image).encuda()
 }
 
+// TODO: parse jit options
 #[cfg_attr(not(test), no_mangle)]
 pub extern "C" fn cuModuleLoadDataEx(
     module: *mut CUmodule,
@@ -2462,7 +2463,7 @@ pub extern "C" fn cuModuleLoadDataEx(
     options: *mut CUjit_option,
     optionValues: *mut *mut ::std::os::raw::c_void,
 ) -> CUresult {
-    r#impl::unimplemented()
+    r#impl::module::load_data(module.decuda(), image).encuda()
 }
 
 #[cfg_attr(not(test), no_mangle)]
@@ -2736,6 +2737,16 @@ pub extern "C" fn cuMemcpyHtoD_v2(
     r#impl::memory::copy_v2(dstDevice.decuda(), srcHost, ByteCount).encuda()
 }
 
+// TODO: implement default stream semantics
+#[cfg_attr(not(test), no_mangle)]
+pub extern "C" fn cuMemcpyHtoD_v2_ptds(
+    dstDevice: CUdeviceptr,
+    srcHost: *const ::std::os::raw::c_void,
+    ByteCount: usize,
+) -> CUresult {
+    r#impl::memory::copy_v2(dstDevice.decuda(), srcHost, ByteCount).encuda()
+}
+
 #[cfg_attr(not(test), no_mangle)]
 pub extern "C" fn cuMemcpyDtoH_v2(
     dstHost: *mut ::std::os::raw::c_void,
@@ -2745,6 +2756,16 @@ pub extern "C" fn cuMemcpyDtoH_v2(
     r#impl::memory::copy_v2(dstHost, srcDevice.decuda(), ByteCount).encuda()
 }
 
+// TODO: implement default stream semantics
+#[cfg_attr(not(test), no_mangle)]
+pub extern "C" fn cuMemcpyDtoH_v2_ptds(
+    dstHost: *mut ::std::os::raw::c_void,
+    srcDevice: CUdeviceptr,
+    ByteCount: usize,
+) -> CUresult {
+    r#impl::memory::copy_v2(dstHost, srcDevice.decuda(), ByteCount).encuda()
+}
+
 #[cfg_attr(not(test), no_mangle)]
 pub extern "C" fn cuMemcpyDtoD_v2(
     dstDevice: CUdeviceptr,
@@ -2926,6 +2947,16 @@ pub extern "C" fn cuMemsetD8_v2(
     r#impl::memory::set_d8_v2(dstDevice.decuda(), uc, N).encuda()
 }
 
+// TODO: implement default stream semantics
+#[cfg_attr(not(test), no_mangle)]
+pub extern "C" fn cuMemsetD8_v2_ptds(
+    dstDevice: CUdeviceptr,
+    uc: ::std::os::raw::c_uchar,
+    N: usize,
+) -> CUresult {
+    r#impl::memory::set_d8_v2(dstDevice.decuda(), uc, N).encuda()
+}
+
 #[cfg_attr(not(test), no_mangle)]
 pub extern "C" fn cuMemsetD16_v2(
     dstDevice: CUdeviceptr,
@@ -2944,6 +2975,16 @@ pub extern "C" fn cuMemsetD32_v2(
     r#impl::memory::set_d32_v2(dstDevice.decuda(), ui, N).encuda()
 }
 
+// TODO: implement default stream semantics
+#[cfg_attr(not(test), no_mangle)]
+pub extern "C" fn cuMemsetD32_v2_ptds(
+    dstDevice: CUdeviceptr,
+    ui: ::std::os::raw::c_uint,
+    N: usize,
+) -> CUresult {
+    r#impl::memory::set_d32_v2(dstDevice.decuda(), ui, N).encuda()
+}
+
 #[cfg_attr(not(test), no_mangle)]
 pub extern "C" fn cuMemsetD2D8_v2(
     dstDevice: CUdeviceptr,
@@ -3322,6 +3363,12 @@ pub extern "C" fn cuStreamGetCtx(hStream: CUstream, pctx: *mut CUcontext) -> CUr
     r#impl::stream::get_ctx(hStream.decuda(), pctx.decuda()).encuda()
 }
 
+// TODO: implement default stream semantics
+#[cfg_attr(not(test), no_mangle)]
+pub extern "C" fn cuStreamGetCtx_ptsz(hStream: CUstream, pctx: *mut CUcontext) -> CUresult {
+    r#impl::stream::get_ctx(hStream.decuda(), pctx.decuda()).encuda()
+}
+
 #[cfg_attr(not(test), no_mangle)]
 pub extern "C" fn cuStreamWaitEvent(
     hStream: CUstream,
@@ -3630,6 +3677,37 @@ pub extern "C" fn cuLaunchKernel(
     .encuda()
 }
 
+// TODO: implement default stream semantics
+#[cfg_attr(not(test), no_mangle)]
+pub extern "C" fn cuLaunchKernel_ptsz(
+    f: CUfunction,
+    gridDimX: ::std::os::raw::c_uint,
+    gridDimY: ::std::os::raw::c_uint,
+    gridDimZ: ::std::os::raw::c_uint,
+    blockDimX: ::std::os::raw::c_uint,
+    blockDimY: ::std::os::raw::c_uint,
+    blockDimZ: ::std::os::raw::c_uint,
+    sharedMemBytes: ::std::os::raw::c_uint,
+    hStream: CUstream,
+    kernelParams: *mut *mut ::std::os::raw::c_void,
+    extra: *mut *mut ::std::os::raw::c_void,
+) -> CUresult {
+    r#impl::function::launch_kernel(
+        f.decuda(),
+        gridDimX,
+        gridDimY,
+        gridDimZ,
+        blockDimX,
+        blockDimY,
+        blockDimZ,
+        sharedMemBytes,
+        hStream.decuda(),
+        kernelParams,
+        extra,
+    )
+    .encuda()
+}
+
 #[cfg_attr(not(test), no_mangle)]
 pub extern "C" fn cuLaunchCooperativeKernel(
     f: CUfunction,
diff --git a/zluda/src/impl/module.rs b/zluda/src/impl/module.rs
index bdfcd86..98580f8 100644
--- a/zluda/src/impl/module.rs
+++ b/zluda/src/impl/module.rs
@@ -87,7 +87,7 @@ impl SpirvModule {
         };
         let l0_module = match self.should_link_ptx_impl {
             None => {
-                l0::Module::build_spirv(ctx, dev, byte_il, Some(self.build_options.as_c_str())).0
+                l0::Module::build_spirv(ctx, dev, byte_il, Some(self.build_options.as_c_str()))
             }
             Some(ptx_impl) => {
                 l0::Module::build_link_spirv(