From 90801467d8b0f71e516c960dc6f0f24ac1aa3405 Mon Sep 17 00:00:00 2001 From: Angelos Katharopoulos Date: Tue, 4 Mar 2025 16:21:15 -0800 Subject: [PATCH] Stop the fence in the destructor --- mlx/backend/metal/fence.cpp | 6 ++++++ mlx/backend/metal/fence.h | 1 + 2 files changed, 7 insertions(+) diff --git a/mlx/backend/metal/fence.cpp b/mlx/backend/metal/fence.cpp index 54ce8ea8d..949ce7e23 100644 --- a/mlx/backend/metal/fence.cpp +++ b/mlx/backend/metal/fence.cpp @@ -34,6 +34,12 @@ Fence::Fence(const Stream& stream) : stream_(stream) { } } +Fence::~Fence() { + if (use_fast_) { + cpu_value()[0] = INT_MAX; + } +} + void Fence::wait_gpu(array& x) { gpu_count_++; auto& d = metal::device(stream_.device); diff --git a/mlx/backend/metal/fence.h b/mlx/backend/metal/fence.h index 28e04e643..7d53d469e 100644 --- a/mlx/backend/metal/fence.h +++ b/mlx/backend/metal/fence.h @@ -21,6 +21,7 @@ namespace mlx::core { class Fence { public: Fence(const Stream& stream); + ~Fence(); void update_gpu(const array& x); void wait_gpu(array& x);