Add empty cublas

Cargo.lock

@@ -1343,6 +1343,22 @@ dependencies = [
  "syn 2.0.89",
 ]
 
+[[package]]
+name = "zluda_blas"
+version = "0.0.0"
+dependencies = [
+ "cuda_base",
+ "cuda_types",
+]
+
+[[package]]
+name = "zluda_blaslt"
+version = "0.0.0"
+
+[[package]]
+name = "zluda_dnn"
+version = "0.0.0"
+
 [[package]]
 name = "zluda_dump"
 version = "0.0.0"
@@ -1364,6 +1380,10 @@ dependencies = [
  "winapi",
 ]
 
+[[package]]
+name = "zluda_fft"
+version = "0.0.0"
+
 [[package]]
 name = "zluda_inject"
 version = "0.0.0"

Cargo.toml

@@ -20,6 +20,10 @@ members = [
     "ptx_parser_macros_impl",
     "xtask",
     "zluda_bindgen",
+    "zluda_dnn",
+    "zluda_blas",
+    "zluda_blaslt",
+    "zluda_fft",
 ]
 
 default-members = ["zluda", "zluda_ml", "zluda_inject", "zluda_redirect"]

cuda_base/src/cublaslt.rs

@@ -107,7 +107,7 @@ extern "system" {
         type_: cuda_types::cublaslt::cudaDataType,
         rows: cuda_types::cublaslt::u64,
         cols: cuda_types::cublaslt::u64,
-        ld: cuda_types::cublaslt::i64,
+        ld: i64,
     ) -> cuda_types::cublaslt::cublasStatus_t;
     #[must_use]
     /** Create new matrix layout descriptor.
@@ -119,7 +119,7 @@ extern "system" {
         type_: cuda_types::cublaslt::cudaDataType,
         rows: cuda_types::cublaslt::u64,
         cols: cuda_types::cublaslt::u64,
-        ld: cuda_types::cublaslt::i64,
+        ld: i64,
     ) -> cuda_types::cublaslt::cublasStatus_t;
     #[must_use]
     /** Destroy matrix layout descriptor.

cuda_base/src/cudnn8.rs

@@ -2558,7 +2558,7 @@ extern "system" {
         descriptor: cuda_types::cudnn8::cudnnBackendDescriptor_t,
         attributeName: cuda_types::cudnn8::cudnnBackendAttributeName_t,
         attributeType: cuda_types::cudnn8::cudnnBackendAttributeType_t,
-        elementCount: cuda_types::cudnn8::i64,
+        elementCount: i64,
         arrayOfElements: *const ::core::ffi::c_void,
     ) -> cuda_types::cudnn8::cudnnStatus_t;
     #[must_use]
@@ -2566,8 +2566,8 @@ extern "system" {
         descriptor: cuda_types::cudnn8::cudnnBackendDescriptor_t,
         attributeName: cuda_types::cudnn8::cudnnBackendAttributeName_t,
         attributeType: cuda_types::cudnn8::cudnnBackendAttributeType_t,
-        requestedElementCount: cuda_types::cudnn8::i64,
-        elementCount: *mut cuda_types::cudnn8::i64,
+        requestedElementCount: i64,
+        elementCount: *mut i64,
         arrayOfElements: *mut ::core::ffi::c_void,
     ) -> cuda_types::cudnn8::cudnnStatus_t;
     #[must_use]

cuda_base/src/cudnn9.rs

@@ -75,7 +75,7 @@ extern "system" {
         descriptor: cuda_types::cudnn9::cudnnBackendDescriptor_t,
         attributeName: cuda_types::cudnn9::cudnnBackendAttributeName_t,
         attributeType: cuda_types::cudnn9::cudnnBackendAttributeType_t,
-        elementCount: cuda_types::cudnn9::i64,
+        elementCount: i64,
         arrayOfElements: *const ::core::ffi::c_void,
     ) -> cuda_types::cudnn9::cudnnStatus_t;
     #[must_use]
@@ -83,8 +83,8 @@ extern "system" {
         descriptor: cuda_types::cudnn9::cudnnBackendDescriptor_t,
         attributeName: cuda_types::cudnn9::cudnnBackendAttributeName_t,
         attributeType: cuda_types::cudnn9::cudnnBackendAttributeType_t,
-        requestedElementCount: cuda_types::cudnn9::i64,
-        elementCount: *mut cuda_types::cudnn9::i64,
+        requestedElementCount: i64,
+        elementCount: *mut i64,
         arrayOfElements: *mut ::core::ffi::c_void,
     ) -> cuda_types::cudnn9::cudnnStatus_t;
     #[must_use]

cuda_base/src/lib.rs

@@ -15,6 +15,8 @@ use syn::{
 
 const CUDA_RS: &'static str = include_str! {"cuda.rs"};
 const NVML_RS: &'static str = include_str! {"nvml.rs"};
+const CUBLAS_RS: &'static str = include_str! {"cublas.rs"};
+const CUBLASLT_RS: &'static str = include_str! {"cublaslt.rs"};
 
 // This macro accepts following arguments:
 // * `normal_macro`: ident for a normal macro
@@ -35,6 +37,16 @@ pub fn cuda_function_declarations(tokens: TokenStream) -> TokenStream {
     function_declarations(tokens, CUDA_RS)
 }
 
+#[proc_macro]
+pub fn cublas_function_declarations(tokens: TokenStream) -> TokenStream {
+    function_declarations(tokens, CUBLAS_RS)
+}
+
+#[proc_macro]
+pub fn cublaslt_function_declarations(tokens: TokenStream) -> TokenStream {
+    function_declarations(tokens, CUBLASLT_RS)
+}
+
 fn function_declarations(tokens: TokenStream, module: &str) -> TokenStream {
     let input = parse_macro_input!(tokens as FnDeclInput);
     let mut cuda_module = syn::parse_str::<File>(module).unwrap();

cuda_types/src/cublas.rs

@@ -1,6 +1,14 @@
 // Generated automatically by zluda_bindgen
 // DO NOT EDIT MANUALLY
 #![allow(warnings)]
+pub type __half = u16;
+pub type __nv_bfloat16 = u16;
+pub use super::cuda::cuComplex;
+pub use super::cuda::cuDoubleComplex;
+pub use super::cuda::cudaDataType;
+pub use super::cuda::cudaDataType_t;
+pub type cudaStream_t = super::cuda::CUstream;
+pub use super::cuda::libraryPropertyType;
 pub const CUBLAS_VER_MAJOR: u32 = 12;
 pub const CUBLAS_VER_MINOR: u32 = 8;
 pub const CUBLAS_VER_PATCH: u32 = 4;

cuda_types/src/cublaslt.rs

@@ -1,6 +1,14 @@
 // Generated automatically by zluda_bindgen
 // DO NOT EDIT MANUALLY
 #![allow(warnings)]
+pub type __half = u16;
+pub type __nv_bfloat16 = u16;
+pub use super::cuda::cuComplex;
+pub use super::cuda::cuDoubleComplex;
+pub use super::cuda::cudaDataType;
+pub use super::cuda::cudaDataType_t;
+pub type cudaStream_t = super::cuda::CUstream;
+pub use super::cuda::libraryPropertyType;
 pub const CUBLASLT_NUMERICAL_IMPL_FLAGS_FMA: u32 = 1;
 pub const CUBLASLT_NUMERICAL_IMPL_FLAGS_HMMA: u32 = 2;
 pub const CUBLASLT_NUMERICAL_IMPL_FLAGS_IMMA: u32 = 4;

cuda_types/src/cuda.rs

@@ -8423,6 +8423,19 @@ impl cudaDataType_t {
 #[derive(Debug, Copy, Clone, Hash, PartialEq, Eq)]
 pub struct cudaDataType_t(pub ::core::ffi::c_uint);
 pub use self::cudaDataType_t as cudaDataType;
+impl libraryPropertyType_t {
+    pub const MAJOR_VERSION: libraryPropertyType_t = libraryPropertyType_t(0);
+}
+impl libraryPropertyType_t {
+    pub const MINOR_VERSION: libraryPropertyType_t = libraryPropertyType_t(1);
+}
+impl libraryPropertyType_t {
+    pub const PATCH_LEVEL: libraryPropertyType_t = libraryPropertyType_t(2);
+}
+#[repr(transparent)]
+#[derive(Debug, Copy, Clone, Hash, PartialEq, Eq)]
+pub struct libraryPropertyType_t(pub ::core::ffi::c_uint);
+pub use self::libraryPropertyType_t as libraryPropertyType;
 #[repr(C)]
 #[repr(align(8))]
 #[derive(Debug, Copy, Clone, PartialEq)]

cuda_types/src/cudnn8.rs

@@ -1,6 +1,14 @@
 // Generated automatically by zluda_bindgen
 // DO NOT EDIT MANUALLY
 #![allow(warnings)]
+pub type __half = u16;
+pub type __nv_bfloat16 = u16;
+pub use super::cuda::cuComplex;
+pub use super::cuda::cuDoubleComplex;
+pub use super::cuda::cudaDataType;
+pub use super::cuda::cudaDataType_t;
+pub type cudaStream_t = super::cuda::CUstream;
+pub use super::cuda::libraryPropertyType;
 pub const CUDNN_MAJOR: u32 = 8;
 pub const CUDNN_MINOR: u32 = 9;
 pub const CUDNN_PATCHLEVEL: u32 = 7;
@@ -196,7 +204,7 @@ pub struct cudnnDebugStruct {
     pub time_usec: ::core::ffi::c_uint,
     pub time_delta: ::core::ffi::c_uint,
     pub handle: cudnnHandle_t,
-    pub stream: super::cuda::CUstream,
+    pub stream: cudaStream_t,
     pub pid: ::core::ffi::c_ulonglong,
     pub tid: ::core::ffi::c_ulonglong,
     pub cudaDeviceId: ::core::ffi::c_int,

cuda_types/src/cudnn9.rs

@@ -1,6 +1,14 @@
 // Generated automatically by zluda_bindgen
 // DO NOT EDIT MANUALLY
 #![allow(warnings)]
+pub type __half = u16;
+pub type __nv_bfloat16 = u16;
+pub use super::cuda::cuComplex;
+pub use super::cuda::cuDoubleComplex;
+pub use super::cuda::cudaDataType;
+pub use super::cuda::cudaDataType_t;
+pub type cudaStream_t = super::cuda::CUstream;
+pub use super::cuda::libraryPropertyType;
 pub const CUDNN_MAJOR: u32 = 9;
 pub const CUDNN_MINOR: u32 = 8;
 pub const CUDNN_PATCHLEVEL: u32 = 0;
@@ -330,7 +338,7 @@ pub struct cudnnDebugStruct {
     pub time_usec: ::core::ffi::c_uint,
     pub time_delta: ::core::ffi::c_uint,
     pub handle: cudnnHandle_t,
-    pub stream: super::cuda::CUstream,
+    pub stream: cudaStream_t,
     pub pid: ::core::ffi::c_ulonglong,
     pub tid: ::core::ffi::c_ulonglong,
     pub cudaDeviceId: ::core::ffi::c_int,

cuda_types/src/cufft.rs

@@ -1,6 +1,14 @@
 // Generated automatically by zluda_bindgen
 // DO NOT EDIT MANUALLY
 #![allow(warnings)]
+pub type __half = u16;
+pub type __nv_bfloat16 = u16;
+pub use super::cuda::cuComplex;
+pub use super::cuda::cuDoubleComplex;
+pub use super::cuda::cudaDataType;
+pub use super::cuda::cudaDataType_t;
+pub type cudaStream_t = super::cuda::CUstream;
+pub use super::cuda::libraryPropertyType;
 pub const CUFFT_VER_MAJOR: u32 = 11;
 pub const CUFFT_VER_MINOR: u32 = 3;
 pub const CUFFT_VER_PATCH: u32 = 3;

cuda_types/src/nvml.rs

@@ -1,6 +1,14 @@
 // Generated automatically by zluda_bindgen
 // DO NOT EDIT MANUALLY
 #![allow(warnings)]
+pub type __half = u16;
+pub type __nv_bfloat16 = u16;
+pub use super::cuda::cuComplex;
+pub use super::cuda::cuDoubleComplex;
+pub use super::cuda::cudaDataType;
+pub use super::cuda::cudaDataType_t;
+pub type cudaStream_t = super::cuda::CUstream;
+pub use super::cuda::libraryPropertyType;
 pub const NVML_API_VERSION: u32 = 12;
 pub const NVML_API_VERSION_STR: &[u8; 3] = b"12\0";
 pub const NVML_VALUE_NOT_AVAILABLE: i32 = -1;

zluda_bindgen/src/main.rs

@@ -179,7 +179,7 @@ fn write_cudnn9_types(
     };
     let mut output = output.clone();
     output.extend(cudnn9_path);
-    let text = prettyplease::unparse(&module).replace("cudaStream_t", "super::cuda::CUstream");
+    let text = prettyplease::unparse(&module);
     write_rust_to_file(output, &text)
 }
 
@@ -241,7 +241,7 @@ fn write_cudnn8_types(
     };
     let mut output = output.clone();
     output.extend(cudnn8_path);
-    let text = prettyplease::unparse(&module).replace("cudaStream_t", "super::cuda::CUstream");
+    let text = prettyplease::unparse(&module);
     write_rust_to_file(output, &text)
 }
 
@@ -487,6 +487,7 @@ fn generate_cuda(crate_root: &PathBuf) {
         .allowlist_type("^CU.*")
         .allowlist_type("^cuda.*")
         .allowlist_type("^cu.*Complex.*")
+        .allowlist_type("^libraryPropertyType.*")
         .allowlist_function("^cu.*")
         .allowlist_var("^CU.*")
         .must_use_type("cudaError_enum")
@@ -574,18 +575,29 @@ fn generate_types_library(crate_root: &PathBuf, path: &[&str], module: &syn::Fil
     let text = prettyplease::unparse(&module)
         .replace("self::cudaDataType", "super::cuda::cudaDataType")
         // complex as used by cuFFT
-        .replace("cuComplex", "super::cuda::cuComplex")
-        .replace("cuDoubleComplex", "super::cuda::cuDoubleComplex");
+        .replace(" cuComplex", " super::cuda::cuComplex")
+        .replace(" cuDoubleComplex", " super::cuda::cuDoubleComplex");
     write_rust_to_file(output, &text)
 }
 
 fn generate_types_library_impl(module: &syn::File) -> syn::File {
+    let known_reexports: Punctuated<syn::Item, syn::parse::Nothing> = parse_quote! {
+        pub type __half = u16;
+        pub type __nv_bfloat16 = u16;
+        pub use super::cuda::cuComplex;
+        pub use super::cuda::cuDoubleComplex;
+        pub use super::cuda::cudaDataType;
+        pub use super::cuda::cudaDataType_t;
+        pub type cudaStream_t = super::cuda::CUstream;
+        pub use super::cuda::libraryPropertyType;
+    };
     let non_fn = module.items.iter().filter_map(|item| match item {
         Item::ForeignMod(_) => None,
         _ => Some(item),
     });
+    let items = known_reexports.iter().chain(non_fn);
     parse_quote! {
-            #(#non_fn)*
+        #(#items)*
     }
 }
 
@@ -851,7 +863,7 @@ impl VisitMut for PrependCudaPath {
     fn visit_type_path_mut(&mut self, type_: &mut TypePath) {
         if type_.path.segments.len() == 1 {
             match &*type_.path.segments[0].ident.to_string() {
-                "usize" | "f64" | "f32" => {}
+                "usize" | "i64" | "f64" | "f32" => {}
                 _ => {
                     let module = &self.module;
                     *type_ = parse_quote! { cuda_types :: #module :: #type_ };
@@ -907,6 +919,7 @@ fn generate_display(
         "CUdevResource_st",
         "CUlaunchAttribute_st",
         "CUlaunchConfig_st",
+        "CUmemcpy3DOperand_st",
     ];
     let ignore_functions = [
         "cuGLGetDevices",

zluda_blas/Cargo.toml

@@ -0,0 +1,17 @@
+[package]
+name = "zluda_blas"
+version = "0.0.0"
+edition = "2021"
+
+[lib]
+crate-type = ["cdylib"]
+name = "cublas"
+
+[dependencies]
+cuda_base = { path = "../cuda_base" }
+cuda_types = { path = "../cuda_types" }
+
+[package.metadata.zluda]
+linux_symlinks = [
+    "libcublas.so.12",
+]

zluda_blas/src/impl.rs

@@ -0,0 +1,24 @@
+use cuda_types::cublas::cublasStatus_t;
+
+#[cfg(debug_assertions)]
+pub(crate) fn unimplemented() -> cublasStatus_t {
+    unimplemented!()
+}
+
+#[cfg(not(debug_assertions))]
+pub(crate) fn unimplemented() -> cublasStatus_t {
+    CUresult::ERROR_NOT_SUPPORTED
+}
+
+pub fn cublasGetStatusName(status: cuda_types::cublas::cublasStatus_t) -> *const ::core::ffi::c_char {
+    todo!()
+}
+pub fn cublasGetStatusString(status: cuda_types::cublas::cublasStatus_t) -> *const ::core::ffi::c_char {
+    todo!()
+}
+pub fn cublasXerbla(srName: *const ::core::ffi::c_char, info: ::core::ffi::c_int) -> () {
+    todo!()
+}
+pub fn cublasGetCudartVersion() -> usize {
+    todo!()
+}

zluda_blas/src/lib.rs

@@ -0,0 +1,37 @@
+mod r#impl;
+
+macro_rules! unimplemented {
+    ($($abi:literal fn $fn_name:ident( $($arg_id:ident : $arg_type:ty),* ) -> $ret_type:ty;)*) => {
+        $(
+            #[cfg_attr(not(test), no_mangle)]
+            #[allow(improper_ctypes)]
+            #[allow(improper_ctypes_definitions)]
+            pub unsafe extern $abi fn $fn_name ( $( $arg_id : $arg_type),* ) -> $ret_type {
+                crate::r#impl::unimplemented()
+            }
+        )*
+    };
+}
+
+macro_rules! implemented {
+    ($($abi:literal fn $fn_name:ident( $($arg_id:ident : $arg_type:ty),* ) -> $ret_type:ty;)*) => {
+        $(
+            #[cfg_attr(not(test), no_mangle)]
+            #[allow(improper_ctypes)]
+            #[allow(improper_ctypes_definitions)]
+            pub unsafe extern $abi fn $fn_name ( $( $arg_id : $arg_type),* ) -> $ret_type {
+                crate::r#impl::$fn_name( $( $arg_id ),* )
+            }
+        )*
+    };
+}
+
+cuda_base::cublas_function_declarations!(
+    unimplemented,
+    implemented <= [
+        cublasGetStatusName,
+        cublasGetStatusString,
+        cublasXerbla,
+        cublasGetCudartVersion
+    ]
+);

zluda_blaslt/Cargo.toml

@@ -0,0 +1,16 @@
+[package]
+name = "zluda_blaslt"
+version = "0.0.0"
+edition = "2021"
+
+[lib]
+crate-type = ["cdylib"]
+name = "cublaslt"
+
+[dependencies]
+
+[package.metadata.zluda]
+linux_symlinks = [
+    "libcublasLt.so",
+    "libcublasLt.so.12",
+]

zluda_blaslt/src/lib.rs

@@ -0,0 +1,3 @@
+fn main() {
+    println!("Hello, world!");
+}

zluda_dnn/Cargo.toml

@@ -0,0 +1,16 @@
+[package]
+name = "zluda_dnn"
+version = "0.0.0"
+edition = "2021"
+
+[lib]
+crate-type = ["cdylib"]
+name = "cudnn64_9"
+
+[dependencies]
+
+[package.metadata.zluda]
+linux_symlinks = [
+    "libcudnn.so",
+    "libcudnn.so.9",
+]

zluda_dnn/src/lib.rs

@@ -0,0 +1,3 @@
+fn main() {
+    println!("Hello, world!");
+}

zluda_dump/src/format.rs

@@ -114,6 +114,17 @@ impl CudaDisplay for f32 {
     }
 }
 
+impl CudaDisplay for f64 {
+    fn write(
+        &self,
+        _fn_name: &'static str,
+        _index: usize,
+        writer: &mut (impl std::io::Write + ?Sized),
+    ) -> std::io::Result<()> {
+        write!(writer, "{}", *self)
+    }
+}
+
 pub fn write_handle<T: LowerHex>(
     this: &[T; 64],
     writer: &mut (impl std::io::Write + ?Sized),
@@ -257,6 +268,69 @@ impl CudaDisplay for CUstreamBatchMemOpParams {
     }
 }
 
+impl CudaDisplay for CUcheckpointRestoreArgs_st {
+    fn write(
+        &self,
+        fn_name: &'static str,
+        index: usize,
+        writer: &mut (impl std::io::Write + ?Sized),
+    ) -> std::io::Result<()> {
+        CudaDisplay::write(&self.reserved, fn_name, index, writer)
+    }
+}
+
+impl CudaDisplay for CUcheckpointUnlockArgs_st {
+    fn write(
+        &self,
+        fn_name: &'static str,
+        index: usize,
+        writer: &mut (impl std::io::Write + ?Sized),
+    ) -> std::io::Result<()> {
+        CudaDisplay::write(&self.reserved, fn_name, index, writer)
+    }
+}
+
+impl CudaDisplay for CUcheckpointCheckpointArgs_st {
+    fn write(
+        &self,
+        fn_name: &'static str,
+        index: usize,
+        writer: &mut (impl std::io::Write + ?Sized),
+    ) -> std::io::Result<()> {
+        CudaDisplay::write(&self.reserved, fn_name, index, writer)
+    }
+}
+
+impl CudaDisplay for CUmemcpy3DOperand_st {
+    fn write(
+        &self,
+        fn_name: &'static str,
+        index: usize,
+        writer: &mut (impl std::io::Write + ?Sized),
+    ) -> std::io::Result<()> {
+        writer.write_all(b"{ type_: ")?;
+        CudaDisplay::write(&self.type_, "", 0, writer)?;
+        writer.write_all(b", op: ")?;
+        match self.type_ {
+            CUmemcpy3DOperandType::CU_MEMCPY_OPERAND_TYPE_ARRAY => {
+                CudaDisplay::write(unsafe { &self.op.array }, fn_name, index, writer)?;
+            }
+            CUmemcpy3DOperandType::CU_MEMCPY_OPERAND_TYPE_POINTER => {
+                CudaDisplay::write(unsafe { &self.op.ptr }, fn_name, index, writer)?;
+            }
+            _ => {
+                CudaDisplay::write(
+                    &unsafe { mem::transmute::<_, [u8; 32]>(self.op) },
+                    fn_name,
+                    index,
+                    writer,
+                )?;
+            }
+        }
+        writer.write_all(b" }")
+    }
+}
+
 pub fn write_wait_value(
     this: &CUstreamBatchMemOpParams_union_CUstreamMemOpWaitValueParams_st,
     writer: &mut (impl std::io::Write + ?Sized),
@@ -347,29 +421,29 @@ impl CudaDisplay for CUDA_EXTERNAL_MEMORY_HANDLE_DESC_st {
         match self.type_ {
             CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD => {
                 writer.write_all(b", handle: ")?;
-                CudaDisplay::write(unsafe { &self.handle.fd }, "", 0,writer)?;
+                CudaDisplay::write(unsafe { &self.handle.fd }, "", 0, writer)?;
             }
             CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32
             | CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_D3D12_HEAP
             | CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_D3D12_RESOURCE
-            |CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_D3D11_RESOURCE => {
+            | CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_D3D11_RESOURCE => {
                 write_win32_handle(unsafe { self.handle.win32 }, writer)?;
             }
             CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT
             | CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_D3D11_RESOURCE_KMT => {
                 writer.write_all(b", handle: ")?;
-                CudaDisplay::write(unsafe { &self.handle.win32.handle }, "", 0,writer)?;
+                CudaDisplay::write(unsafe { &self.handle.win32.handle }, "", 0, writer)?;
             }
             CUexternalMemoryHandleType::CU_EXTERNAL_MEMORY_HANDLE_TYPE_NVSCIBUF => {
                 writer.write_all(b", handle: ")?;
-                CudaDisplay::write(unsafe { &self.handle.nvSciBufObject }, "", 0,writer)?;
+                CudaDisplay::write(unsafe { &self.handle.nvSciBufObject }, "", 0, writer)?;
             }
             _ => {
                 writer.write_all(b", size: ")?;
-                CudaDisplay::write(&self.size, "", 0,writer)?;
+                CudaDisplay::write(&self.size, "", 0, writer)?;
                 writer.write_all(b", flags: ")?;
-                CudaDisplay::write(&self.flags, "", 0,writer)?;
-                return writer.write_all(b", ... }")
+                CudaDisplay::write(&self.flags, "", 0, writer)?;
+                return writer.write_all(b", ... }");
             }
         }
         writer.write_all(b", size: ")?;
@@ -441,9 +515,7 @@ impl CudaDisplay for CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC_st {
     }
 }
 
-impl CudaDisplay
-    for CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_st__bindgen_ty_1__bindgen_ty_2
-{
+impl CudaDisplay for CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_st__bindgen_ty_1__bindgen_ty_2 {
     fn write(
         &self,
         _fn_name: &'static str,
@@ -456,9 +528,7 @@ impl CudaDisplay
     }
 }
 
-impl CudaDisplay
-    for CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_st__bindgen_ty_1__bindgen_ty_2
-{
+impl CudaDisplay for CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_st__bindgen_ty_1__bindgen_ty_2 {
     fn write(
         &self,
         _fn_name: &'static str,
@@ -667,15 +737,30 @@ fn write_launch_attribute(
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_CLUSTER_SCHEDULING_POLICY_PREFERENCE => {
             writer.write_all(b", value_out: ")?;
-            CudaDisplay::write(unsafe { &(*value_out).clusterSchedulingPolicyPreference }, fn_name, index, writer)
+            CudaDisplay::write(
+                unsafe { &(*value_out).clusterSchedulingPolicyPreference },
+                fn_name,
+                index,
+                writer,
+            )
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_PROGRAMMATIC_STREAM_SERIALIZATION => {
             writer.write_all(b", value_out: ")?;
-            CudaDisplay::write(unsafe { &(*value_out).programmaticStreamSerializationAllowed }, fn_name, index, writer)
+            CudaDisplay::write(
+                unsafe { &(*value_out).programmaticStreamSerializationAllowed },
+                fn_name,
+                index,
+                writer,
+            )
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_PROGRAMMATIC_EVENT => {
             writer.write_all(b", value_out: ")?;
-            CudaDisplay::write(unsafe { &(*value_out).programmaticEvent }, fn_name, index, writer)
+            CudaDisplay::write(
+                unsafe { &(*value_out).programmaticEvent },
+                fn_name,
+                index,
+                writer,
+            )
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_PRIORITY => {
             writer.write_all(b", value_out: ")?;
@@ -683,19 +768,39 @@ fn write_launch_attribute(
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_MEM_SYNC_DOMAIN_MAP => {
             writer.write_all(b", value_out: ")?;
-            CudaDisplay::write(unsafe { &(*value_out).memSyncDomainMap }, fn_name, index, writer)
+            CudaDisplay::write(
+                unsafe { &(*value_out).memSyncDomainMap },
+                fn_name,
+                index,
+                writer,
+            )
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_MEM_SYNC_DOMAIN => {
             writer.write_all(b", value_out: ")?;
-            CudaDisplay::write(unsafe { &(*value_out).memSyncDomain }, fn_name, index, writer)
+            CudaDisplay::write(
+                unsafe { &(*value_out).memSyncDomain },
+                fn_name,
+                index,
+                writer,
+            )
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_LAUNCH_COMPLETION_EVENT => {
             writer.write_all(b", value_out: ")?;
-            CudaDisplay::write(unsafe { &(*value_out).launchCompletionEvent }, fn_name, index, writer)
+            CudaDisplay::write(
+                unsafe { &(*value_out).launchCompletionEvent },
+                fn_name,
+                index,
+                writer,
+            )
         }
         CUlaunchAttributeID::CU_LAUNCH_ATTRIBUTE_DEVICE_UPDATABLE_KERNEL_NODE => {
             writer.write_all(b", value_out: ")?;
-            CudaDisplay::write(unsafe { &(*value_out).deviceUpdatableKernelNode }, fn_name, index, writer)
+            CudaDisplay::write(
+                unsafe { &(*value_out).deviceUpdatableKernelNode },
+                fn_name,
+                index,
+                writer,
+            )
         }
         _ => writer.write_all(b", ... "),
     }

zluda_dump/src/format_generated.rs

@@ -6177,20 +6177,6 @@ impl crate::format::CudaDisplay for cuda_types::cuda::CUextent3D_st {
         writer.write_all(b" }")
     }
 }
-impl crate::format::CudaDisplay for cuda_types::cuda::CUmemcpy3DOperand_st {
-    fn write(
-        &self,
-        _fn_name: &'static str,
-        _index: usize,
-        writer: &mut (impl std::io::Write + ?Sized),
-    ) -> std::io::Result<()> {
-        writer.write_all(concat!("{ ", stringify!(type_), ": ").as_bytes())?;
-        crate::format::CudaDisplay::write(&self.type_, "", 0, writer)?;
-        writer.write_all(concat!(", ", stringify!(op), ": ").as_bytes())?;
-        crate::format::CudaDisplay::write(&self.op, "", 0, writer)?;
-        writer.write_all(b" }")
-    }
-}
 impl crate::format::CudaDisplay
 for cuda_types::cuda::CUmemcpy3DOperand_st__bindgen_ty_1__bindgen_ty_1 {
     fn write(
@@ -26370,6 +26356,27 @@ impl crate::format::CudaDisplay for cuda_types::cuda::cudaDataType_t {
         }
     }
 }
+impl crate::format::CudaDisplay for cuda_types::cuda::libraryPropertyType_t {
+    fn write(
+        &self,
+        _fn_name: &'static str,
+        _index: usize,
+        writer: &mut (impl std::io::Write + ?Sized),
+    ) -> std::io::Result<()> {
+        match self {
+            &cuda_types::cuda::libraryPropertyType_t::MAJOR_VERSION => {
+                writer.write_all(stringify!(MAJOR_VERSION).as_bytes())
+            }
+            &cuda_types::cuda::libraryPropertyType_t::MINOR_VERSION => {
+                writer.write_all(stringify!(MINOR_VERSION).as_bytes())
+            }
+            &cuda_types::cuda::libraryPropertyType_t::PATCH_LEVEL => {
+                writer.write_all(stringify!(PATCH_LEVEL).as_bytes())
+            }
+            _ => write!(writer, "{}", self.0),
+        }
+    }
+}
 impl crate::format::CudaDisplay for cuda_types::cuda::float2 {
     fn write(
         &self,

zluda_fft/Cargo.toml

@@ -0,0 +1,15 @@
+[package]
+name = "zluda_fft"
+version = "0.0.0"
+edition = "2021"
+
+[lib]
+crate-type = ["cdylib"]
+name = "cufft"
+
+[dependencies]
+
+[package.metadata.zluda]
+linux_symlinks = [
+    "libcufft.so.11",
+]

zluda_fft/src/lib.rs

@@ -0,0 +1,3 @@
+fn main() {
+    println!("Hello, world!");
+}