Skip to content

Correct the inline macros #21

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
adamdempsey90 merged 1 commit into from
Jul 29, 2026
Merged

Correct the inline macros #21

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
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
14 changes: 7 additions & 7 deletions pips/device/device_common.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,17 +1,17 @@
#ifndef PIPS_DEVICE_DEVICE_COMMON_HPP_
#define PIPS_DEVICE_DEVICE_COMMON_HPP_

// Qualifier used on functions that must be callable from both host code and
// CUDA/HIP device code. On plain C++ compilers it expands to nothing.
// Qualifiers for functions that must be callable from both host code and
// CUDA/HIP device code.
#if defined(__CUDACC__) || defined(__HIPCC__)
#define PIPS_DEVICE_HOST __host__ __device__ inline
#define PIPS_DEVICE_HOST_INLINE PIPS_DEVICE_HOST __forceinline__
#define PIPS_DEVICE_HOST_INLINE __host__ __device__ inline
#define PIPS_DEVICE_HOST_FORCEINLINE __host__ __device__ __forceinline__
#else
#define PIPS_DEVICE_HOST inline
#define PIPS_DEVICE_HOST_INLINE inline
#if defined(__GNUC__) || defined(__clang__)
#define PIPS_DEVICE_HOST_INLINE __attribute__((always_inline)) inline
#define PIPS_DEVICE_HOST_FORCEINLINE __attribute__((always_inline)) inline
#else
#define PIPS_DEVICE_HOST_INLINE PIPS_DEVICE_HOST
#define PIPS_DEVICE_HOST_FORCEINLINE inline
#endif
#endif

Expand Down
28 changes: 14 additions & 14 deletions pips/device/device_value.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -45,28 +45,28 @@ struct DeviceValue {
} as;
};

PIPS_DEVICE_HOST_INLINE DeviceValue dv_nil() {
PIPS_DEVICE_HOST_FORCEINLINE DeviceValue dv_nil() {
DeviceValue v{};
v.type = DeviceValueType::NIL;
v.as.n = 0;
return v;
}

PIPS_DEVICE_HOST_INLINE DeviceValue dv_bool(bool x) {
PIPS_DEVICE_HOST_FORCEINLINE DeviceValue dv_bool(bool x) {
DeviceValue v{};
v.type = DeviceValueType::BOOL;
v.as.b = x;
return v;
}

PIPS_DEVICE_HOST_INLINE DeviceValue dv_number(DeviceReal x) {
PIPS_DEVICE_HOST_FORCEINLINE DeviceValue dv_number(DeviceReal x) {
DeviceValue v{};
v.type = DeviceValueType::NUMBER;
v.as.n = x;
return v;
}

PIPS_DEVICE_HOST_INLINE bool dv_vector(const DeviceReal *elements,
PIPS_DEVICE_HOST_FORCEINLINE bool dv_vector(const DeviceReal *elements,
std::uint32_t length,
DeviceValue &out) {
if (length > PIPS_DEVICE_VECTOR_MAX || (length > 0 && !elements))
Expand All @@ -80,37 +80,37 @@ PIPS_DEVICE_HOST_INLINE bool dv_vector(const DeviceReal *elements,
return true;
}

PIPS_DEVICE_HOST_INLINE constexpr bool dv_is_nil(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr bool dv_is_nil(const DeviceValue &v) {
return v.type == DeviceValueType::NIL;
}
PIPS_DEVICE_HOST_INLINE constexpr bool dv_is_bool(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr bool dv_is_bool(const DeviceValue &v) {
return v.type == DeviceValueType::BOOL;
}
PIPS_DEVICE_HOST_INLINE constexpr bool dv_is_number(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr bool dv_is_number(const DeviceValue &v) {
return v.type == DeviceValueType::NUMBER;
}
PIPS_DEVICE_HOST_INLINE constexpr bool dv_is_vector(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr bool dv_is_vector(const DeviceValue &v) {
return v.type == DeviceValueType::VECTOR;
}
PIPS_DEVICE_HOST_INLINE constexpr bool dv_as_bool(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr bool dv_as_bool(const DeviceValue &v) {
return v.as.b;
}
PIPS_DEVICE_HOST_INLINE constexpr DeviceReal dv_as_number(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr DeviceReal dv_as_number(const DeviceValue &v) {
return v.as.n;
}
PIPS_DEVICE_HOST_INLINE constexpr std::uint8_t
PIPS_DEVICE_HOST_FORCEINLINE constexpr std::uint8_t
dv_vector_length(const DeviceValue &v) {
return v.as.vector.length;
}
PIPS_DEVICE_HOST_INLINE constexpr DeviceReal
PIPS_DEVICE_HOST_FORCEINLINE constexpr DeviceReal
dv_vector_element(const DeviceValue &v, std::uint8_t index) {
return v.as.vector.elements[index];
}
PIPS_DEVICE_HOST_INLINE constexpr bool dv_vector_is_valid(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr bool dv_vector_is_valid(const DeviceValue &v) {
return !dv_is_vector(v) || v.as.vector.length <= PIPS_DEVICE_VECTOR_MAX;
}

PIPS_DEVICE_HOST_INLINE constexpr bool dv_is_falsey(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE constexpr bool dv_is_falsey(const DeviceValue &v) {
return dv_is_nil(v) || (dv_is_bool(v) && !v.as.b);
}

Expand Down
14 changes: 7 additions & 7 deletions pips/device/device_vm.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -34,18 +34,18 @@ struct DeviceVM {
std::int32_t sp = 0;
std::int32_t fp = 0;

PIPS_DEVICE_HOST_INLINE bool push(const DeviceValue &v) {
PIPS_DEVICE_HOST_FORCEINLINE bool push(const DeviceValue &v) {
if (sp >= PIPS_DEVICE_STACK_MAX) return false;
stack[sp++] = v;
return true;
}
PIPS_DEVICE_HOST_INLINE DeviceValue pop() { return stack[--sp]; }
PIPS_DEVICE_HOST_INLINE const DeviceValue &peek(int back) const {
PIPS_DEVICE_HOST_FORCEINLINE DeviceValue pop() { return stack[--sp]; }
PIPS_DEVICE_HOST_FORCEINLINE const DeviceValue &peek(int back) const {
return stack[sp - 1 - back];
}

// Execute a packed function
PIPS_DEVICE_HOST DeviceStatus run(const DeviceModule &module,
PIPS_DEVICE_HOST_INLINE DeviceStatus run(const DeviceModule &module,
std::uint32_t entry_id,
const DeviceValue *args,
std::uint32_t argc,
Expand All @@ -72,7 +72,7 @@ struct DeviceVM {
}

private:
PIPS_DEVICE_HOST_INLINE static DeviceStatus
PIPS_DEVICE_HOST_FORCEINLINE static DeviceStatus
scalar_arith(DeviceOpCode op, DeviceReal a, DeviceReal b, DeviceReal &out) {
switch (op) {
case DeviceOpCode::ADD: out = a + b; return DeviceStatus::OK;
Expand All @@ -93,7 +93,7 @@ struct DeviceVM {
}
}

PIPS_DEVICE_HOST_INLINE static DeviceStatus
PIPS_DEVICE_HOST_FORCEINLINE static DeviceStatus
vector_arith(DeviceOpCode op, const DeviceValue &a, const DeviceValue &b,
DeviceValue &out) {
const bool av = dv_is_vector(a);
Expand Down Expand Up @@ -128,7 +128,7 @@ struct DeviceVM {
return DeviceStatus::OK;
}

PIPS_DEVICE_HOST DeviceStatus dispatch(const DeviceModule &module,
PIPS_DEVICE_HOST_INLINE DeviceStatus dispatch(const DeviceModule &module,
DeviceValue *out_result) {
using OC = DeviceOpCode;

Expand Down
Loading