Skip to content

Commit e1a6bc5

Browse files
bweltonclaude
andauthored
[rocprofiler-sdk] Fix queues_init leak in on-demand queue mode (#3683)
Replace function-static queues_init set with a static_object accessor function and erase entries when on-demand queues are destroyed. Previously, each start/stop cycle created a new queue pointer that accumulated in the set, causing unbounded growth and dangling pointer correctness issues if HSA reused an address. Co-authored-by: Claude Opus 4.6 <noreply@anthropic.com>
1 parent 1d06e29 commit e1a6bc5

1 file changed

Lines changed: 13 additions & 3 deletions

File tree

projects/rocprofiler-sdk/source/lib/rocprofiler-sdk/counters/device_counting.cpp

Lines changed: 13 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -23,6 +23,7 @@
2323
#include "lib/rocprofiler-sdk/counters/device_counting.hpp"
2424
#include "lib/common/environment.hpp"
2525
#include "lib/common/logging.hpp"
26+
#include "lib/common/static_object.hpp"
2627
#include "lib/common/utility.hpp"
2728
#include "lib/rocprofiler-sdk/buffer.hpp"
2829
#include "lib/rocprofiler-sdk/context/context.hpp"
@@ -95,6 +96,15 @@ submitPacket(hsa_queue_t* queue, const void* packet)
9596

9697
namespace
9798
{
99+
// Tracks which queues have had set_profiler_active_on_queue called.
100+
// Accessor function avoids static initialization order issues.
101+
std::unordered_set<hsa_queue_t*>&
102+
get_queues_init()
103+
{
104+
static auto& _v = *common::static_object<std::unordered_set<hsa_queue_t*>>::construct();
105+
return _v;
106+
}
107+
98108
// Returns true if device lock should be acquired at configuration time (OLD behavior).
99109
// Returns false if device lock should be acquired at context start time (NEW behavior, default).
100110
bool
@@ -260,9 +270,8 @@ init_callback_data(rocprofiler::counters::agent_callback_data& callback_data,
260270

261271
// If we do not have a completion handle, this is our first time profiling this agent.
262272
// Setup our shared data structures.
263-
static std::unordered_set<hsa_queue_t*> queues_init;
264-
if(queues_init.find(callback_data.queue) != queues_init.end()) return;
265-
queues_init.insert(callback_data.queue);
273+
if(get_queues_init().find(callback_data.queue) != get_queues_init().end()) return;
274+
get_queues_init().insert(callback_data.queue);
266275

267276
// Set state of the queue to allow profiling (may not be needed since AQL
268277
// may do this in the future).
@@ -633,6 +642,7 @@ stop_agent_ctx(const context::context* ctx)
633642
hsa::get_core_table()->hsa_signal_destroy_fn(callback_data.start_signal);
634643
callback_data.start_signal.handle = 0;
635644
}
645+
get_queues_init().erase(callback_data.queue);
636646
callback_data.packet.reset();
637647
callback_data.queue = nullptr;
638648
agent->destroy_device_counting_service_queue();

0 commit comments

Comments
 (0)