Skip to content

Add fallback definition for metal::isnan #1463

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
merged 1 commit into from
Jun 11, 2025
Merged

Add fallback definition for metal::isnan #1463

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
9 changes: 9 additions & 0 deletions mistralrs-quant/src/metal_kernels/utils.metal
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -263,6 +263,15 @@ inline bfloat16_t uint16_to_bfloat16(const uint16_t x) {
return _MLX_BFloat16(x, _MLX_BFloat16::bits_to_bfloat());
}

// Resolve ambiguous call to metal::isnan for bfloat16_t by providing an
// exact‑match overload. This prevents the compiler from having to choose
// between the existing float and half overloads when the argument is a
// bfloat16_t value.
METAL_FUNC inline bool isnan(bfloat16_t x) {
// Delegate to the float overload after an explicit cast.
return metal::isnan(static_cast<float>(x));
}

Comment on lines +266 to +274
Copy link

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

⚠️ Potential issue

Ensure the isnan overload is declared in the metal namespace
The new overload is currently defined in the global namespace (via using namespace metal), so qualified calls to metal::isnan(bfloat16_t) won’t pick it up. Move or wrap this overload in namespace metal so it properly overloads metal::isnan for bfloat16_t.

Proposed diff:

-// Resolve ambiguous call to metal::isnan for bfloat16_t by providing an
-// exact-match overload.  This prevents the compiler from having to choose
-// between the existing float and half overloads when the argument is a
-// bfloat16_t value.
-METAL_FUNC inline bool isnan(bfloat16_t x) {
-  // Delegate to the float overload after an explicit cast.
-  return metal::isnan(static_cast<float>(x));
-}
+namespace metal {
+// Resolve ambiguous call to metal::isnan for bfloat16_t by providing an
+// exact-match overload. This prevents ambiguity between float and half.
+METAL_FUNC inline bool isnan(bfloat16_t x) {
+  // Delegate to the float overload after an explicit cast.
+  return ::metal::isnan(static_cast<float>(x));
+}
+} // namespace metal
📝 Committable suggestion

‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.

Suggested change
// Resolve ambiguous call to metal::isnan for bfloat16_t by providing an
// exact‑match overload. This prevents the compiler from having to choose
// between the existing float and half overloads when the argument is a
// bfloat16_t value.
METAL_FUNC inline bool isnan(bfloat16_t x) {
// Delegate to the float overload after an explicit cast.
return metal::isnan(static_cast<float>(x));
}
namespace metal {
// Resolve ambiguous call to metal::isnan for bfloat16_t by providing an
// exact-match overload. This prevents ambiguity between float and half.
METAL_FUNC inline bool isnan(bfloat16_t x) {
// Delegate to the float overload after an explicit cast.
return ::metal::isnan(static_cast<float>(x));
}
} // namespace metal
🤖 Prompt for AI Agents 
In mistralrs-quant/src/metal_kernels/utils.metal around lines 266 to 274, the
isnan overload for bfloat16_t is defined in the global namespace, which prevents
metal::isnan(bfloat16_t) from resolving correctly. To fix this, move or wrap the
isnan function inside the metal namespace so it properly overloads metal::isnan
for bfloat16_t, ensuring qualified calls find this overload.

#endif

///////////////////////////////////////////////////////////////////////////////
Expand Down
Loading