On 2026-11-09 16:35, Han, Qiang wrote:
AMD General
cc @Zhao, Huasha @Gu, Zhenyu
-----Original Message-----
From: Yang, Philip <[email protected]>
Sent: Monday, November 9, 2026 1:15 PM
To: [email protected]; Kuehling, Felix <[email protected]>; Gandham,
Shashidhar <[email protected]>; Han, Qiang <[email protected]>
Cc: Han, Qiang <[email protected]>; Yang, Philip <[email protected]>
Subject: [PATCH] drm/amdkfd: wait for the queue quiesce in concurrent SVM
evictions
From: Qiang Han <[email protected]>
With XNACK off, svm_range_evict() leaves the GPU page tables in place and
relies on the process's queues being stopped before the MMU notifier returns.
Only the first eviction (evicted_ranges == 1) calls kgd2kfd_quiesce_mm(). The
quiesce preempts every GPU the process has opened and takes milliseconds. Any
other range invalidated in the meantime sees evicted_ranges > 1 and returns at
once. The core mm then frees, migrates or reuses the page while the queues still
run, and an in-flight SDMA or blit copy accesses it through the old GPU PTE. The
result is silent corruption of the copy, or of whatever process now owns the page.
Hold a new svms->evict_lock across the first evictor's quiesce, so that concurrent
evictions return only after the queues have stopped. This is the scheme
amdgpu_amdkfd_evict_userptr() uses with process_info->notifier_lock. evict_lock nests
inside prange->lock and is taken nowhere else. Under it, the quiesce path takes the
dqm lock, which is already used in MMU notifiers (memalloc_noreclaim).
One way to reach this: a pageable hipMemcpy() larger than 1 MiB. ROCclr pins
the buffer with hsa_amd_memory_lock(), libhsakmt registers it as a KFD SVM
range, and host memory pressure (reclaim, compaction, allocator purge)
invalidates the range.
Fixes: 8a7c184a16d7 ("drm/amdkfd: svm range eviction and restore")
Signed-off-by: Qing Hang <[email protected]>
Signed-off-by: Philip Yang <[email protected]>
Reviewed-by: Felix Kuehling <[email protected]>
---
drivers/gpu/drm/amd/amdkfd/kfd_priv.h | 2 ++
drivers/gpu/drm/amd/amdkfd/kfd_svm.c | 35 ++++++++++++++++++---------
2 files changed, 25 insertions(+), 12 deletions(-)
diff --git a/drivers/gpu/drm/amd/amdkfd/kfd_priv.h
b/drivers/gpu/drm/amd/amdkfd/kfd_priv.h
index 158fa70a1a8e..581fb7ceba40 100644
--- a/drivers/gpu/drm/amd/amdkfd/kfd_priv.h
+++ b/drivers/gpu/drm/amd/amdkfd/kfd_priv.h
@@ -909,6 +909,8 @@ struct svm_range_list {
struct list_head deferred_range_list;
struct list_head criu_svm_metadata_list;
spinlock_t deferred_list_lock;
+ /* serializes XNACK-off evictions with the queue quiesce */
+ struct mutex evict_lock;
atomic_t evicted_ranges;
atomic_t drain_pagefaults;
struct delayed_work restore_work;
diff --git a/drivers/gpu/drm/amd/amdkfd/kfd_svm.c
b/drivers/gpu/drm/amd/amdkfd/kfd_svm.c
index 3688ae7ba35e..8b84333a85b7 100644
--- a/drivers/gpu/drm/amd/amdkfd/kfd_svm.c
+++ b/drivers/gpu/drm/amd/amdkfd/kfd_svm.c
@@ -2141,21 +2141,30 @@ svm_range_evict(struct svm_range *prange, struct
mm_struct *mm,
if (prange->start <= last && prange->last >= start)
atomic_inc(&prange->invalid);
+ /* The MMU notifier must not return before the GPU has stopped
+ * accessing the invalidated pages. Only the first eviction
stops
+ * the queues, which takes milliseconds (HWS preemption on every
+ * GPU of the process). Serialize all evictions of this process
so
+ * that a concurrent invalidation of another range waits until
+ * that quiesce has completed instead of returning while queues
+ * still run on stale PTEs.
+ */
+ mutex_lock(&svms->evict_lock);
evicted_ranges = atomic_inc_return(&svms->evicted_ranges);
- if (evicted_ranges != 1)
- return r;
-
- pr_debug("evicting svms 0x%p range [0x%lx 0x%lx]\n",
- prange->svms, prange->start, prange->last);
+ if (evicted_ranges == 1) {
+ pr_debug("evicting svms 0x%p range [0x%lx 0x%lx]\n",
+ prange->svms, prange->start, prange->last);
- /* First eviction, stop the queues */
- r = kgd2kfd_quiesce_mm(mm, KFD_QUEUE_EVICTION_TRIGGER_SVM);
- if (r)
- pr_debug("failed to quiesce KFD\n");
+ /* First eviction, stop the queues */
+ r = kgd2kfd_quiesce_mm(mm,
KFD_QUEUE_EVICTION_TRIGGER_SVM);
+ if (r)
+ pr_debug("failed to quiesce KFD\n");
- pr_debug("schedule to restore svm %p ranges\n", svms);
- queue_delayed_work(system_freezable_wq, &svms->restore_work,
- msecs_to_jiffies(AMDGPU_SVM_RANGE_RESTORE_DELAY_MS));
+ pr_debug("schedule to restore svm %p ranges\n", svms);
+ queue_delayed_work(system_freezable_wq,
&svms->restore_work,
+
msecs_to_jiffies(AMDGPU_SVM_RANGE_RESTORE_DELAY_MS));
+ }
+ mutex_unlock(&svms->evict_lock);
} else {
unsigned long s, l;
uint32_t trigger;
@@ -3437,6 +3446,7 @@ void svm_range_list_fini(struct kfd_process *p)
svm_range_free(prange, true);
}
+ mutex_destroy(&p->svms.evict_lock);
mutex_destroy(&p->svms.lock);
pr_debug("process pid %d svms 0x%p done\n", @@ -3450,6 +3460,7 @@ int
svm_range_list_init(struct kfd_process *p)
svms->objects = RB_ROOT_CACHED;
mutex_init(&svms->lock);
+ mutex_init(&svms->evict_lock);
INIT_LIST_HEAD(&svms->list);
atomic_set(&svms->evicted_ranges, 0);
atomic_set(&svms->drain_pagefaults, 0);
--
2.50.1
