Is cuda_std still maintained?

[Jark5455]

Hi, I am currently working on my own cuda bindings for rust. I was looking for some implementations for basic math functions in cuda kernels, and I came across this project. I tried compiling cuda_std targeting nvptx but I got a lot of compiler errors related to missing imports + removed features.

Is cuda_std still maintained? I noticed that the last changes were 4+ years ago. Is there something else yall are using now?

[Y-jiji]

I think it is maintained. Can you see whether you can run the following code?

lab6.zip

[LegNeato]

Yes? What errors are you hitting?

[Jark5455]

Sorry for the late response and overall shitty issue, should have gave more details, I was busy with finals, cuda_std doesnt appear to compile with the latest nightly build

Compiling cuda_std v0.2.2
error[E0557]: feature has been removed
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/lib.rs:28:9
              |
           28 |         register_attr,
              |         ^^^^^^^^^^^^^ feature has been removed
              |
              = note: removed in 1.65.0; see <https://github.com/rust-lang/rust/pull/66070> for more information
              = note: removed in favor of `#![register_tool]`
           
           error: environment variable `CUDA_ARCH` not defined at compile time
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/cfg.rs:32:15
              |
           32 |         match env!("CUDA_ARCH") {
              |               ^^^^^^^^^^^^^^^^^
              |
              = help: use `std::env::var("CUDA_ARCH")` to read the variable at run time
           
           error: cannot find attribute `register_attr` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/lib.rs:34:5
              |
           34 |     register_attr(nvvm_internal)
              |     ^^^^^^^^^^^^^
              |
           help: a built-in attribute with a similar name exists
              |
           34 -     register_attr(nvvm_internal)
           34 +     register_tool(nvvm_internal)
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/mem.rs:40:9
              |
           40 |         asm!(
              |         ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/mem.rs:58:9
              |
           58 |         asm!(
              |         ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/misc.rs:10:9
              |
           10 |         asm!("brkpt");
              |         ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/misc.rs:28:9
              |
           28 |         asm!(
              |         ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/misc.rs:41:9
              |
           41 |         asm!(
              |         ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
            --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:9:5
             |
           9 |     asm!("membar.gl;");
             |     ^^^
             |
           help: consider importing this macro
             |
           3 + use core::arch::asm;
             |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:14:5
              |
           14 |     asm!("membar.cta;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:19:5
              |
           19 |     asm!("membar.sys;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:24:5
              |
           24 |     asm!("fence.sc.gl;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:29:5
              |
           29 |     asm!("fence.sc.cta;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:34:5
              |
           34 |     asm!("fence.sc.sys;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:39:5
              |
           39 |     asm!("fence.acq_rel.gl;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:44:5
              |
           44 |     asm!("fence.acq_rel.sys;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/atomic/intrinsics.rs:49:5
              |
           49 |     asm!("fence.acq_rel.sys;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:30:5
              |
           30 |     asm!(".reg .pred p;");
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:34:33
              |
           34 |         AddressSpace::Global => asm!("isspacep.global p, {}", in(reg64) ptr),
              |                                 ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:35:33
              |
           35 |         AddressSpace::Shared => asm!("isspacep.shared p, {}", in(reg64) ptr),
              |                                 ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:36:35
              |
           36 |         AddressSpace::Constant => asm!("isspacep.const p, {}", in(reg64) ptr),
              |                                   ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:37:32
              |
           37 |         AddressSpace::Local => asm!("isspacep.local p, {}", in(reg64) ptr),
              |                                ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:41:5
              |
           41 |     asm!("selp.u32 {}, 1, 0, p;", out(reg32) ret);
              |     ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:61:33
              |
           61 |         AddressSpace::Global => asm!(
              |                                 ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:66:33
              |
           66 |         AddressSpace::Shared => asm!(
              |                                 ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:71:35
              |
           71 |         AddressSpace::Constant => asm!(
              |                                   ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:76:32
              |
           76 |         AddressSpace::Local => asm!(
              |                                ^^^
              |
           help: consider importing this macro
              |
            3 + use core::arch::asm;
              |
           
           error: cannot find macro `asm` in this scope
              --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:101:33
               |
           101 |         AddressSpace::Global => asm!(
               |                                 ^^^
               |
           help: consider importing this macro
               |
             3 + use core::arch::asm;
               |
           
           error: cannot find macro `asm` in this scope
              --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:106:33
               |
           106 |         AddressSpace::Shared => asm!(
               |                                 ^^^
               |
           help: consider importing this macro
               |
             3 + use core::arch::asm;
               |
           
           error: cannot find macro `asm` in this scope
              --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:111:35
               |
           111 |         AddressSpace::Constant => asm!(
               |                                   ^^^
               |
           help: consider importing this macro
               |
             3 + use core::arch::asm;
               |
           
           error: cannot find macro `asm` in this scope
              --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/ptr.rs:116:32
               |
           116 |         AddressSpace::Local => asm!(
               |                                ^^^
               |
           help: consider importing this macro
               |
             3 + use core::arch::asm;
               |
           
           error: cannot find attribute `nvvm_internal` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/shared.rs:80:11
              |
           80 |         #[nvvm_internal(addrspace(3))]
              |           ^^^^^^^^^^^^^
           
           error: cannot find macro `asm` in this scope
              --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/thread.rs:347:9
               |
           347 |         asm!(
               |         ^^^
               |
           help: consider importing this macro
               |
            21 + use core::arch::asm;
               |
           
           error: cannot find macro `asm` in this scope
             --> /home/yashr/.cargo/registry/src/index.crates.io-1949cf8c6b5b557f/cuda_std-0.2.2/src/warp.rs:36:9
              |
           36 |         asm!(
              |         ^^^
              |
           help: consider importing this macro
              |
            6 + use core::arch::asm;
              |
           
           error[E0152]: duplicate lang item in crate `core` (which `alloc` depends on): `sized`
             |
             = note: the lang item is first defined in crate `core` (which `cuda_std` depends on)
             = note: first definition in `core` loaded from /home/yashr/RustroverProjects/cuda_std_test/target/nvptx64-nvidia-cuda/release/deps/libcore-c11c987fb9d840c3.rmeta
             = note: second definition in `core` loaded from /home/yashr/.rustup/toolchains/nightly-x86_64-unknown-linux-gnu/lib/rustlib/nvptx64-nvidia-cuda/lib/libcore-2c3ca0f476147ac0.rmeta
           
           Some errors have detailed explanations: E0152, E0557.
           For more information about an error, try `rustc --explain E0152`.
           error: could not compile `cuda_std` (lib) due to 35 previous errors; 4 warnings emitted

Am I required to use the nvvm backend with it?

[LegNeato]

No worries on the issue, it happens. What nightly? The project requires the rustc version in https://github.com/Rust-GPU/rust-cuda/blob/main/rust-toolchain.toml to work as it is tied closely with internals. You might try using our docker images: https://github.com/orgs/Rust-GPU/packages?repo_name=rust-cuda

[LegNeato]

cuda_std works with the upstream ptx64-nvidia-cuda target, though there are a couple of places where it has dependencies on the nvvm backend where they rendevous (but if you don't use those, it will work fine).