Skip to content
Merged
Show file tree
Hide file tree
Changes from 3 commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
3 changes: 2 additions & 1 deletion docs/cccl/development/macro.rst
Original file line number Diff line number Diff line change
Expand Up @@ -294,7 +294,8 @@ Usage example:
+----------------------------------+------------------------------------------------------------------------------+
| ``_CCCL_CONST`` | Portable "constant" function attribute |
+----------------------------------+------------------------------------------------------------------------------+

| ``_CCCL_LIFETIMEBOUND`` | Portable "lifetime bound" function attribute |
+----------------------------------+------------------------------------------------------------------------------+

**Portable Builtin Macros**:

Expand Down
12 changes: 12 additions & 0 deletions libcudacxx/include/cuda/std/__cccl/attributes.h
Original file line number Diff line number Diff line change
Expand Up @@ -154,6 +154,18 @@
#define _CCCL_NO_SPECIALIZATIONS \
_CCCL_NO_SPECIALIZATIONS_BECAUSE("Users are not allowed to specialize this cccl entity")

// _CCCL_LIFETIMEBOUND

#if _CCCL_HAS_CPP_ATTRIBUTE(clang::lifetimebound) || _CCCL_COMPILER(CLANG)
# define _CCCL_LIFETIMEBOUND [[clang::lifetimebound]]
#elif _CCCL_HAS_CPP_ATTRIBUTE(msvc::lifetimebound) && _CCCL_COMPILER(MSVC, >=, 19, 37)
# define _CCCL_LIFETIMEBOUND [[msvc::lifetimebound]]
#elif _CCCL_HAS_CPP_ATTRIBUTE(lifetimebound)
# define _CCCL_LIFETIMEBOUND [[lifetimebound]]
#else
# define _CCCL_LIFETIMEBOUND
#endif

// _CCCL_NO_UNIQUE_ADDRESS

#if _CCCL_COMPILER(MSVC) || _CCCL_HAS_CPP_ATTRIBUTE(no_unique_address) < 201803L
Expand Down
Original file line number Diff line number Diff line change
@@ -0,0 +1,39 @@
//===----------------------------------------------------------------------===//
//
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
// See https://llvm.org/LICENSE.txt for license information.
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
// SPDX-FileCopyrightText: Copyright (c) 2026 NVIDIA CORPORATION & AFFILIATES.
//
//===----------------------------------------------------------------------===//

#include <cuda/std/__cccl/attributes.h>

#include "test_macros.h"

#if (_CCCL_LIFETIMEBOUND + 0)
#else
# error "lifetimebound attribute not supported"
#endif

struct S
{
char data[32];

__host__ __device__ const char* get() const _CCCL_LIFETIMEBOUND
{
return data;
}
};

__host__ __device__ bool test()
{
auto sv = S{"abc"}.get();
return true;
}

int main(int, char**)
{
assert(test());
return 0;
}
Original file line number Diff line number Diff line change
@@ -0,0 +1,25 @@
//===----------------------------------------------------------------------===//
//
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
// See https://llvm.org/LICENSE.txt for license information.
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
// SPDX-FileCopyrightText: Copyright (c) 2026 NVIDIA CORPORATION & AFFILIATES.
//
//===----------------------------------------------------------------------===//

#include <cuda/std/__cccl/attributes.h>

#include "test_macros.h"

struct S
{
__host__ __device__ int get() const _CCCL_LIFETIMEBOUND
{
return 3;
}
};

int main(int, char**)
{
return 0;
}