cuda::bitmask

Contents

cuda::bitmask#

2026-10-07

1 min read time

Applies to Linux and Windows

This page documents cuda::bitmask, which generates an integer bitmask of a specified width starting at a given bit position.

template <typename T>
[[nodiscard]] constexpr T
bitmask(int start, int width) noexcept;

The function generates a bitmask of size width starting at position start.

Parameters

  • start: Starting position of the bitmask.

  • width: Width of the bitmask.

Return value

Bitmask of size width starting at start.

Constraints

T is an unsigned integral type.

Preconditions

  • start >= 0 && start <= num_bits(T)

  • width >= 0 && width <= num_bits(T)

  • start + width <= num_bits(T)

Performance considerations

Note

When the input values are run-time values that the compiler can resolve at compile-time, e.g. an index of a loop with a fixed number of iterations, using the function could not be optimal.

Example#

#include <cuda/bit>
#include <cuda/std/cassert>
#include <cuda/std/cstdint>

__global__ void bitmask_kernel() {
    assert(cuda::bitmask<uint32_t>(2, 4) == 0b111100u);
    assert(cuda::bitmask<uint64_t>(1, 3) == uint64_t{0b1110});
}

int main() {
    bitmask_kernel<<<1, 1>>>();
    hipDeviceSynchronize();
    return 0;
}
Contents