Use perf_event_open for context switch tracking.

2 files changed, 457 insertions, 806 deletions

diff --git a/client/TracySysTrace.cpp b/client/TracySysTrace.cpp
index 1972a345..b8e059fd 100644
--- a/client/TracySysTrace.cpp
+++ b/client/TracySysTrace.cpp
@@ -641,31 +641,13 @@ void SysTraceGetExternalName( uint64_t thread, const char*& threadName, const ch
 #    include "TracyRingBuffer.hpp"
 #    include "TracyThread.hpp"
 
-#    ifdef __ANDROID__
-#      include "TracySysTracePayload.hpp"
-#    endif
-
-#    ifdef __AVX2__
-#      include <immintrin.h>
-#    endif
-
 namespace tracy
 {
 
-static const char BasePath[] = "/sys/kernel/debug/tracing/";
-static const char TracingOn[] = "tracing_on";
-static const char CurrentTracer[] = "current_tracer";
-static const char TraceOptions[] = "trace_options";
-static const char TraceClock[] = "trace_clock";
-static const char SchedSwitch[] = "events/sched/sched_switch/enable";
-static const char SchedWakeup[] = "events/sched/sched_wakeup/enable";
-static const char BufferSizeKb[] = "buffer_size_kb";
-static const char TracePipe[] = "trace_pipe";
-
 static std::atomic<bool> traceActive { false };
-static Thread* s_threadSampling = nullptr;
 static int s_numCpus = 0;
 static int s_numBuffers = 0;
+static int s_ctxBufferIdx = 0;
 
 static constexpr size_t RingBufSize = 64*1024;
 static RingBuffer<RingBufSize>* s_ring = nullptr;
@@ -683,7 +665,9 @@ enum TraceEventId
     EventCacheReference,
     EventCacheMiss,
     EventBranchRetired,
-    EventBranchMiss
+    EventBranchMiss,
+    EventContextSwitch,
+    EventWakeup,
 };
 
 static void ProbePreciseIp( perf_event_attr& pe, unsigned long long config0, unsigned long long config1, pid_t pid )
@@ -745,12 +729,103 @@ static bool IsGenuineIntel()
 #endif
 }
 
-static void SetupSampling( int64_t& samplingPeriod )
+static const char* ReadFile( const char* path )
+{
+    int fd = open( path, O_RDONLY );
+    if( fd < 0 ) return nullptr;
+
+    static char tmp[64];
+    const auto cnt = read( fd, tmp, 63 );
+    close( fd );
+    if( cnt < 0 ) return nullptr;
+    tmp[cnt] = '\0';
+    return tmp;
+}
+
+#ifdef __ANDROID__
+static const char* ReadFileElevated( const char* path )
+{
+    // Explanation for "su root sh -c": there are 2 flavors of "su" in circulation
+    // on Android. The default Android su has the following syntax to run a command
+    // as root:
+    //   su root 'command'
+    // and 'command' is exec'd not passed to a shell, so if shell interpretation is
+    // wanted, one needs to do:
+    //   su root sh -c 'command'
+    // Besides that default Android 'su' command, some Android devices use a different
+    // su with a command-line interface closer to the familiar util-linux su found
+    // on Linux distributions. Fortunately, both the util-linux su and the one
+    // in https://github.com/topjohnwu/Magisk seem to be happy with the above
+    // `su root sh -c 'command'` command line syntax.
+
+    int pipefd[2];
+    if( pipe( pipefd ) == 0 )
+    {
+        const auto pid = fork();
+        if( pid == 0 )
+        {
+            // child
+            close( pipefd[0] );
+            dup2( open( "/dev/null", O_WRONLY ), STDERR_FILENO );
+            if( dup2( pipefd[1], STDOUT_FILENO ) >= 0 )
+            {
+                close( pipefd[1] );
+                char tmp[1024];
+                sprintf( tmp, "cat %s", path );
+                execlp( "su", "su", "root", "sh", "-c", tmp, (char*)nullptr );
+                exit( 1 );
+            }
+            exit( 0 );
+        }
+        else if( pid > 0 )
+        {
+            // parent
+            close( pipefd[1] );
+            static char tmp[64];
+            const auto sz = read( pipefd[0], tmp, 63 );
+            close( pipefd[0] );
+            waitpid( pid, nullptr, 0 );
+            if( sz <= 0 ) return nullptr;
+            tmp[sz] = '\0';
+            return tmp;
+        }
+        else
+        {
+            return nullptr;
+        }
+    }
+    else
+    {
+        return nullptr;
+    }
+}
+#else
+static const char* ReadFileElevated( const char* path )
+{
+    return ReadFile( path );
+}
+#endif
+
+bool SysTraceStart( int64_t& samplingPeriod )
 {
 #ifndef CLOCK_MONOTONIC_RAW
-    return;
+    return false;
 #endif
 
+    int paranoidLevel = 2;
+    const auto paranoidLevelStr = ReadFile( "/proc/sys/kernel/perf_event_paranoid" );
+    if( !paranoidLevelStr ) return false;
+    paranoidLevel = atoi( paranoidLevelStr );
+    TracyDebug( "perf_event_paranoid: %i\n", paranoidLevel );
+
+    int switchId = -1, wakeupId = -1;
+    const auto switchIdStr = ReadFileElevated( "/sys/kernel/debug/tracing/events/sched/sched_switch/id" );
+    if( switchIdStr ) switchId = atoi( switchIdStr );
+    const auto wakeupIdStr = ReadFileElevated( "/sys/kernel/debug/tracing/events/sched/sched_wakeup/id" );
+    if( wakeupIdStr ) wakeupId = atoi( wakeupIdStr );
+
+    TracyDebug( "sched_switch id: %i\nsched_wakeup id: %i\n", switchId, wakeupId );
+
 #ifdef TRACY_NO_SAMPLE_RETIREMENT
     const bool noRetirement = true;
 #else
@@ -772,14 +847,28 @@ static void SetupSampling( int64_t& samplingPeriod )
     const bool noBranch = noBranchEnv && noBranchEnv[0] == '1';
 #endif
 
+#ifdef TRACY_NO_CONTEXT_SWITCH
+    const bool noCtxSwitch = true;
+#else
+    const char* noCtxSwitchEnv = GetEnvVar( "TRACY_NO_CONTEXT_SWITCH" );
+    const bool noCtxSwitch = noCtxSwitchEnv && noCtxSwitchEnv[0] == '1';
+#endif
+
     samplingPeriod = GetSamplingPeriod();
     uint32_t currentPid = (uint32_t)getpid();
 
+    const auto maxNumBuffers = s_numCpus * (
+        1 +     // software sampling
+        2 +     // CPU cycles + instructions retired
+        2 +     // cache reference + miss
+        2 +     // branch retired + miss
+        2       // context switches + wakeups
+    );
     s_numCpus = (int)std::thread::hardware_concurrency();
-    s_ring = (RingBuffer<RingBufSize>*)tracy_malloc( sizeof( RingBuffer<RingBufSize> ) * s_numCpus * 7 );
+    s_ring = (RingBuffer<RingBufSize>*)tracy_malloc( sizeof( RingBuffer<RingBufSize> ) * maxNumBuffers );
     s_numBuffers = 0;
 
-    // Stack traces
+    // software sampling
     perf_event_attr pe = {};
     pe.type = PERF_TYPE_SOFTWARE;
     pe.size = sizeof( perf_event_attr );
@@ -801,15 +890,23 @@ static void SetupSampling( int64_t& samplingPeriod )
     ProbePreciseIp( pe, currentPid );
     for( int i=0; i<s_numCpus; i++ )
     {
-        const int fd = perf_event_open( &pe, currentPid, i, -1, PERF_FLAG_FD_CLOEXEC );
+        int fd = perf_event_open( &pe, currentPid, i, -1, PERF_FLAG_FD_CLOEXEC );
         if( fd == -1 )
         {
-            for( int j=0; j<s_numBuffers; j++ ) s_ring[j].~RingBuffer<RingBufSize>();
-            tracy_free( s_ring );
-            return;
+            pe.exclude_kernel = 1;
+            ProbePreciseIp( pe, currentPid );
+            fd = perf_event_open( &pe, currentPid, i, -1, PERF_FLAG_FD_CLOEXEC );
+            if( fd == -1 )
+            {
+                for( int j=0; j<s_numBuffers; j++ ) s_ring[j].~RingBuffer<RingBufSize>();
+                tracy_free( s_ring );
+                return false;
+            }
+            TracyDebug( "  No access to kernel samples\n" );
         }
         new( s_ring+s_numBuffers ) RingBuffer<RingBufSize>( fd, EventCallstack );
         s_numBuffers++;
+        TracyDebug( "  Core %i ok\n", i );
     }
 
     // CPU cycles + instructions retired
@@ -920,766 +1017,398 @@ static void SetupSampling( int64_t& samplingPeriod )
         }
     }
 
+    s_ctxBufferIdx = s_numBuffers;
 
-    s_threadSampling = (Thread*)tracy_malloc( sizeof( Thread ) );
-    new(s_threadSampling) Thread( [] (void*) {
-        ThreadExitHandler threadExitHandler;
-        SetThreadName( "Tracy Sampling" );
-        InitRpmalloc();
-        sched_param sp = { 5 };
-        pthread_setschedparam( pthread_self(), SCHED_FIFO, &sp );
-        for( int i=0; i<s_numBuffers; i++ ) s_ring[i].Enable();
-        for(;;)
-        {
-            bool hadData = false;
-            for( int i=0; i<s_numBuffers; i++ )
-            {
-                if( !traceActive.load( std::memory_order_relaxed ) ) break;
-                auto& ring = s_ring[i];
-                const auto head = ring.LoadHead();
-                const auto tail = ring.GetTail();
-                if( head == tail ) continue;
-                assert( head > tail );
-                hadData = true;
-
-                const auto end = head - tail;
-                uint64_t pos = 0;
-                while( pos < end )
-                {
-                    perf_event_header hdr;
-                    ring.Read( &hdr, pos, sizeof( perf_event_header ) );
-                    if( hdr.type == PERF_RECORD_SAMPLE )
-                    {
-                        auto offset = pos + sizeof( perf_event_header );
-                        const auto id = ring.GetId();
-                        if( id == EventCallstack )
-                        {
-                            // Layout:
-                            //   u32 pid, tid
-                            //   u64 time
-                            //   u64 cnt
-                            //   u64 ip[cnt]
-
-                            uint32_t tid;
-                            uint64_t t0;
-                            uint64_t cnt;
-
-                            offset += sizeof( uint32_t );
-                            ring.Read( &tid, offset, sizeof( uint32_t ) );
-                            offset += sizeof( uint32_t );
-                            ring.Read( &t0, offset, sizeof( uint64_t ) );
-                            offset += sizeof( uint64_t );
-                            ring.Read( &cnt, offset, sizeof( uint64_t ) );
-                            offset += sizeof( uint64_t );
-
-                            if( cnt > 0 )
-                            {
-#if defined TRACY_HW_TIMER && ( defined __i386 || defined _M_IX86 || defined __x86_64__ || defined _M_X64 )
-                                t0 = ring.ConvertTimeToTsc( t0 );
-#endif
-                                auto trace = (uint64_t*)tracy_malloc_fast( ( 1 + cnt ) * sizeof( uint64_t ) );
-                                ring.Read( trace+1, offset, sizeof( uint64_t ) * cnt );
-
-#if defined __x86_64__ || defined _M_X64
-                                // remove non-canonical pointers
-                                do
-                                {
-                                    const auto test = (int64_t)trace[cnt];
-                                    const auto m1 = test >> 63;
-                                    const auto m2 = test >> 47;
-                                    if( m1 == m2 ) break;
-                                }
-                                while( --cnt > 0 );
-                                for( uint64_t j=1; j<cnt; j++ )
-                                {
-                                    const auto test = (int64_t)trace[j];
-                                    const auto m1 = test >> 63;
-                                    const auto m2 = test >> 47;
-                                    if( m1 != m2 ) trace[j] = 0;
-                                }
-#endif
-
-                                for( uint64_t j=1; j<=cnt; j++ )
-                                {
-                                    if( trace[j] >= (uint64_t)-4095 )       // PERF_CONTEXT_MAX
-                                    {
-                                        memmove( trace+j, trace+j+1, sizeof( uint64_t ) * ( cnt - j ) );
-                                        cnt--;
-                                    }
-                                }
-
-                                memcpy( trace, &cnt, sizeof( uint64_t ) );
-
-                                TracyLfqPrepare( QueueType::CallstackSample );
-                                MemWrite( &item->callstackSampleFat.time, t0 );
-                                MemWrite( &item->callstackSampleFat.thread, tid );
-                                MemWrite( &item->callstackSampleFat.ptr, (uint64_t)trace );
-                                TracyLfqCommit;
-                            }
-                        }
-                        else
-                        {
-                            // Layout:
-                            //   u64 ip
-                            //   u64 time
-
-                            uint64_t ip, t0;
-                            ring.Read( &ip, offset, sizeof( uint64_t ) );
-                            offset += sizeof( uint64_t );
-                            ring.Read( &t0, offset, sizeof( uint64_t ) );
-
-#if defined TRACY_HW_TIMER && ( defined __i386 || defined _M_IX86 || defined __x86_64__ || defined _M_X64 )
-                            t0 = ring.ConvertTimeToTsc( t0 );
+    // context switches
+    if( !noCtxSwitch && switchId != -1 )
+    {
+        pe = {};
+        pe.type = PERF_TYPE_TRACEPOINT;
+        pe.size = sizeof( perf_event_attr );
+        pe.sample_period = 1;
+        pe.sample_type = PERF_SAMPLE_TIME | PERF_SAMPLE_RAW;
+        pe.disabled = 1;
+        pe.inherit = 1;
+        pe.config = switchId;
+#if !defined TRACY_HW_TIMER || !( defined __i386 || defined _M_IX86 || defined __x86_64__ || defined _M_X64 )
+        pe.use_clockid = 1;
+        pe.clockid = CLOCK_MONOTONIC_RAW;
 #endif
-                            QueueType type;
-                            switch( id )
-                            {
-                            case EventCpuCycles:
-                                type = QueueType::HwSampleCpuCycle;
-                                break;
-                            case EventInstructionsRetired:
-                                type = QueueType::HwSampleInstructionRetired;
-                                break;
-                            case EventCacheReference:
-                                type = QueueType::HwSampleCacheReference;
-                                break;
-                            case EventCacheMiss:
-                                type = QueueType::HwSampleCacheMiss;
-                                break;
-                            case EventBranchRetired:
-                                type = QueueType::HwSampleBranchRetired;
-                                break;
-                            case EventBranchMiss:
-                                type = QueueType::HwSampleBranchMiss;
-                                break;
-                            default:
-                                assert( false );
-                                break;
-                            }
 
-                            TracyLfqPrepare( type );
-                            MemWrite( &item->hwSample.ip, ip );
-                            MemWrite( &item->hwSample.time, t0 );
-                            TracyLfqCommit;
-                        }
-                    }
-                    pos += hdr.size;
-                }
-                assert( pos == end );
-                ring.Advance( end );
-            }
-            if( !traceActive.load( std::memory_order_relaxed) ) break;
-            if( !hadData )
+        TracyDebug( "Setup context switch capture\n" );
+        for( int i=0; i<s_numCpus; i++ )
+        {
+            const int fd = perf_event_open( &pe, -1, i, -1, PERF_FLAG_FD_CLOEXEC );
+            if( fd != -1 )
             {
-                std::this_thread::sleep_for( std::chrono::milliseconds( 10 ) );
+                new( s_ring+s_numBuffers ) RingBuffer<RingBufSize>( fd, EventContextSwitch, i );
+                s_numBuffers++;
+                TracyDebug( "  Core %i ok\n", i );
             }
         }
 
-        for( int i=0; i<s_numBuffers; i++ ) s_ring[i].~RingBuffer<RingBufSize>();
-        tracy_free_fast( s_ring );
-    }, nullptr );
-}
-
-#ifdef __ANDROID__
-static bool TraceWrite( const char* path, size_t psz, const char* val, size_t vsz )
-{
-    // Explanation for "su root sh -c": there are 2 flavors of "su" in circulation
-    // on Android. The default Android su has the following syntax to run a command
-    // as root:
-    //   su root 'command'
-    // and 'command' is exec'd not passed to a shell, so if shell interpretation is
-    // wanted, one needs to do:
-    //   su root sh -c 'command'
-    // Besides that default Android 'su' command, some Android devices use a different
-    // su with a command-line interface closer to the familiar util-linux su found
-    // on Linux distributions. Fortunately, both the util-linux su and the one
-    // in https://github.com/topjohnwu/Magisk seem to be happy with the above
-    // `su root sh -c 'command'` command line syntax.
-    char tmp[256];
-    sprintf( tmp, "su root sh -c 'echo \"%s\" > %s%s'", val, BasePath, path );
-    return system( tmp ) == 0;
-}
-#else
-static bool TraceWrite( const char* path, size_t psz, const char* val, size_t vsz )
-{
-    char tmp[256];
-    memcpy( tmp, BasePath, sizeof( BasePath ) - 1 );
-    memcpy( tmp + sizeof( BasePath ) - 1, path, psz );
-
-    int fd = open( tmp, O_WRONLY );
-    if( fd < 0 ) return false;
-
-    for(;;)
-    {
-        ssize_t cnt = write( fd, val, vsz );
-        if( cnt == (ssize_t)vsz )
-        {
-            close( fd );
-            return true;
-        }
-        if( cnt < 0 )
+        if( wakeupId != -1 )
         {
-            close( fd );
-            return false;
-        }
-        vsz -= cnt;
-        val += cnt;
-    }
-}
-#endif
+            pe.config = wakeupId;
 
-#ifdef __ANDROID__
-void SysTraceInjectPayload()
-{
-    int pipefd[2];
-    if( pipe( pipefd ) == 0 )
-    {
-        const auto pid = fork();
-        if( pid == 0 )
-        {
-            // child
-            close( pipefd[1] );
-            if( dup2( pipefd[0], STDIN_FILENO ) >= 0 )
+            TracyDebug( "Setup wakeup capture\n" );
+            for( int i=0; i<s_numCpus; i++ )
             {
-                close( pipefd[0] );
-                execlp( "su", "su", "root", "sh", "-c", "cat > /data/tracy_systrace", (char*)nullptr );
-                exit( 1 );
+                const int fd = perf_event_open( &pe, -1, i, -1, PERF_FLAG_FD_CLOEXEC );
+                if( fd != -1 )
+                {
+                    new( s_ring+s_numBuffers ) RingBuffer<RingBufSize>( fd, EventWakeup, i );
+                    s_numBuffers++;
+                    TracyDebug( "  Core %i ok\n", i );
+                }
             }
         }
-        else if( pid > 0 )
-        {
-            // parent
-            close( pipefd[0] );
-
-#ifdef __aarch64__
-            write( pipefd[1], tracy_systrace_aarch64_data, tracy_systrace_aarch64_size );
-#else
-            write( pipefd[1], tracy_systrace_armv7_data, tracy_systrace_armv7_size );
-#endif
-            close( pipefd[1] );
-            waitpid( pid, nullptr, 0 );
-
-            system( "su root sh -c 'chmod 700 /data/tracy_systrace'" );
-        }
     }
-}
-#endif
 
-bool SysTraceStart( int64_t& samplingPeriod )
-{
-#ifndef CLOCK_MONOTONIC_RAW
-    return false;
-#endif
-
-    if( !TraceWrite( TracingOn, sizeof( TracingOn ), "0", 2 ) ) return false;
-    if( !TraceWrite( CurrentTracer, sizeof( CurrentTracer ), "nop", 4 ) ) return false;
-    TraceWrite( TraceOptions, sizeof( TraceOptions ), "norecord-cmd", 13 );
-    TraceWrite( TraceOptions, sizeof( TraceOptions ), "norecord-tgid", 14 );
-    TraceWrite( TraceOptions, sizeof( TraceOptions ), "noirq-info", 11 );
-    TraceWrite( TraceOptions, sizeof( TraceOptions ), "noannotate", 11 );
-#if defined TRACY_HW_TIMER && ( defined __i386 || defined _M_IX86 || defined __x86_64__ || defined _M_X64 )
-    if( !TraceWrite( TraceClock, sizeof( TraceClock ), "x86-tsc", 8 ) ) return false;
-#else
-    if( !TraceWrite( TraceClock, sizeof( TraceClock ), "mono_raw", 9 ) ) return false;
-#endif
-    if( !TraceWrite( SchedSwitch, sizeof( SchedSwitch ), "1", 2 ) ) return false;
-    if( !TraceWrite( SchedWakeup, sizeof( SchedWakeup ), "1", 2 ) ) return false;
-    if( !TraceWrite( BufferSizeKb, sizeof( BufferSizeKb ), "4096", 5 ) ) return false;
-
-#if defined __ANDROID__ && ( defined __aarch64__ || defined __ARM_ARCH )
-    SysTraceInjectPayload();
-#endif
+    TracyDebug( "Ringbuffers in use: %i\n", s_numBuffers );
 
-    if( !TraceWrite( TracingOn, sizeof( TracingOn ), "1", 2 ) ) return false;
     traceActive.store( true, std::memory_order_relaxed );
-
-    SetupSampling( samplingPeriod );
-
     return true;
 }
 
 void SysTraceStop()
 {
-    TraceWrite( TracingOn, sizeof( TracingOn ), "0", 2 );
     traceActive.store( false, std::memory_order_relaxed );
-    if( s_threadSampling )
-    {
-        s_threadSampling->~Thread();
-        tracy_free( s_threadSampling );
-    }
 }
 
-static uint64_t ReadNumber( const char*& data )
+void SysTraceWorker( void* ptr )
 {
-    auto ptr = data;
-    assert( *ptr >= '0' && *ptr <= '9' );
-    uint64_t val = 0;
+    ThreadExitHandler threadExitHandler;
+    SetThreadName( "Tracy Sampling" );
+    InitRpmalloc();
+    sched_param sp = { 5 };
+    pthread_setschedparam( pthread_self(), SCHED_FIFO, &sp );
+    for( int i=0; i<s_numBuffers; i++ ) s_ring[i].Enable();
     for(;;)
     {
-        uint64_t q;
-        memcpy( &q, ptr, 8 );
-        for( int i=0; i<8; i++ )
+        bool hadData = false;
+        for( int i=0; i<s_ctxBufferIdx; i++ )
         {
-            const uint64_t v = ( q & 0xFF ) - '0';
-            if( v > 9 )
+            if( !traceActive.load( std::memory_order_relaxed ) ) break;
+            auto& ring = s_ring[i];
+            const auto head = ring.LoadHead();
+            const auto tail = ring.GetTail();
+            if( head == tail ) continue;
+            assert( head > tail );
+            hadData = true;
+
+            const auto end = head - tail;
+            uint64_t pos = 0;
+            while( pos < end )
             {
-                data = ptr + i;
-                return val;
-            }
-            val = val * 10 + v;
-            q >>= 8;
-        }
-        ptr += 8;
-    }
-}
+                perf_event_header hdr;
+                ring.Read( &hdr, pos, sizeof( perf_event_header ) );
+                if( hdr.type == PERF_RECORD_SAMPLE )
+                {
+                    auto offset = pos + sizeof( perf_event_header );
+                    const auto id = ring.GetId();
+                    assert( id != EventContextSwitch );
+                    if( id == EventCallstack )
+                    {
+                        // Layout:
+                        //   u32 pid, tid
+                        //   u64 time
+                        //   u64 cnt
+                        //   u64 ip[cnt]
+
+                        uint32_t tid;
+                        uint64_t t0;
+                        uint64_t cnt;
+
+                        offset += sizeof( uint32_t );
+                        ring.Read( &tid, offset, sizeof( uint32_t ) );
+                        offset += sizeof( uint32_t );
+                        ring.Read( &t0, offset, sizeof( uint64_t ) );
+                        offset += sizeof( uint64_t );
+                        ring.Read( &cnt, offset, sizeof( uint64_t ) );
+                        offset += sizeof( uint64_t );
+
+                        if( cnt > 0 )
+                        {
+#if defined TRACY_HW_TIMER && ( defined __i386 || defined _M_IX86 || defined __x86_64__ || defined _M_X64 )
+                            t0 = ring.ConvertTimeToTsc( t0 );
+#endif
+                            auto trace = (uint64_t*)tracy_malloc_fast( ( 1 + cnt ) * sizeof( uint64_t ) );
+                            ring.Read( trace+1, offset, sizeof( uint64_t ) * cnt );
 
-static uint8_t ReadState( char state )
-{
-    switch( state )
-    {
-    case 'D': return 101;
-    case 'I': return 102;
-    case 'R': return 103;
-    case 'S': return 104;
-    case 'T': return 105;
-    case 't': return 106;
-    case 'W': return 107;
-    case 'X': return 108;
-    case 'Z': return 109;
-    default: return 100;
-    }
-}
+#if defined __x86_64__ || defined _M_X64
+                            // remove non-canonical pointers
+                            do
+                            {
+                                const auto test = (int64_t)trace[cnt];
+                                const auto m1 = test >> 63;
+                                const auto m2 = test >> 47;
+                                if( m1 == m2 ) break;
+                            }
+                            while( --cnt > 0 );
+                            for( uint64_t j=1; j<cnt; j++ )
+                            {
+                                const auto test = (int64_t)trace[j];
+                                const auto m1 = test >> 63;
+                                const auto m2 = test >> 47;
+                                if( m1 != m2 ) trace[j] = 0;
+                            }
+#endif
 
-#if defined __ANDROID__ && defined __ANDROID_API__ && __ANDROID_API__ < 18
-/*-
- * Copyright (c) 2011 The NetBSD Foundation, Inc.
- * All rights reserved.
- *
- * This code is derived from software contributed to The NetBSD Foundation
- * by Christos Zoulas.
- *
- * Redistribution and use in source and binary forms, with or without
- * modification, are permitted provided that the following conditions
- * are met:
- * 1. Redistributions of source code must retain the above copyright
- *    notice, this list of conditions and the following disclaimer.
- * 2. Redistributions in binary form must reproduce the above copyright
- *    notice, this list of conditions and the following disclaimer in the
- *    documentation and/or other materials provided with the distribution.
- *
- * THIS SOFTWARE IS PROVIDED BY THE NETBSD FOUNDATION, INC. AND CONTRIBUTORS
- * ``AS IS'' AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
- * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
- * PURPOSE ARE DISCLAIMED.  IN NO EVENT SHALL THE FOUNDATION OR CONTRIBUTORS
- * BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR
- * CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF
- * SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS
- * INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN
- * CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE)
- * ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF THE
- * POSSIBILITY OF SUCH DAMAGE.
- */
-
-ssize_t getdelim(char **buf, size_t *bufsiz, int delimiter, FILE *fp)
-{
-	char *ptr, *eptr;
-
-	if (*buf == NULL || *bufsiz == 0) {
-		*bufsiz = BUFSIZ;
-		if ((*buf = (char*)malloc(*bufsiz)) == NULL)
-			return -1;
-	}
-
-	for (ptr = *buf, eptr = *buf + *bufsiz;;) {
-		int c = fgetc(fp);
-		if (c == -1) {
-			if (feof(fp))
-				return ptr == *buf ? -1 : ptr - *buf;
-			else
-				return -1;
-		}
-		*ptr++ = c;
-		if (c == delimiter) {
-			*ptr = '\0';
-			return ptr - *buf;
-		}
-		if (ptr + 2 >= eptr) {
-			char *nbuf;
-			size_t nbufsiz = *bufsiz * 2;
-			ssize_t d = ptr - *buf;
-			if ((nbuf = (char*)realloc(*buf, nbufsiz)) == NULL)
-				return -1;
-			*buf = nbuf;
-			*bufsiz = nbufsiz;
-			eptr = nbuf + nbufsiz;
-			ptr = nbuf + d;
-		}
-	}
-}
+                            for( uint64_t j=1; j<=cnt; j++ )
+                            {
+                                if( trace[j] >= (uint64_t)-4095 )       // PERF_CONTEXT_MAX
+                                {
+                                    memmove( trace+j, trace+j+1, sizeof( uint64_t ) * ( cnt - j ) );
+                                    cnt--;
+                                }
+                            }
 
-ssize_t getline(char **buf, size_t *bufsiz, FILE *fp)
-{
-	return getdelim(buf, bufsiz, '\n', fp);
-}
-#endif
+                            memcpy( trace, &cnt, sizeof( uint64_t ) );
 
-#ifdef __AVX2__
-static inline void AdvanceTo( const char*& line, char match )
-{
-    __m256i m = _mm256_set1_epi8( match );
-    auto ptr = line;
-    for(;;)
-    {
-        __m256i l = _mm256_loadu_si256( (const __m256i*)ptr );
-        __m256i c = _mm256_cmpeq_epi8( l, m );
-        auto b = uint32_t( _mm256_movemask_epi8( c ) );
-        if( b != 0 )
-        {
-            line = ptr + __builtin_ctz( b );
-            return;
-        }
-        ptr += 32;
-    }
-}
-#else
-static inline void AdvanceTo( const char*& line, char match )
-{
-    auto ptr = line;
-    for(;;)
-    {
-        uint64_t l;
-        memcpy( &l, ptr, 8 );
-        for( int i=0; i<8; i++ )
-        {
-            if( ( l & 0xFF ) == uint8_t( match ) )
-            {
-                line = ptr + i;
-                return;
-            }
-            l >>= 8;
-        }
-        ptr += 8;
-    }
-}
-#endif
+                            TracyLfqPrepare( QueueType::CallstackSample );
+                            MemWrite( &item->callstackSampleFat.time, t0 );
+                            MemWrite( &item->callstackSampleFat.thread, tid );
+                            MemWrite( &item->callstackSampleFat.ptr, (uint64_t)trace );
+                            TracyLfqCommit;
+                        }
+                    }
+                    else
+                    {
+                        // Layout:
+                        //   u64 ip
+                        //   u64 time
 
-#ifdef __AVX2__
-static inline void AdvanceToNot( const char*& line, char match )
-{
-    __m256i m = _mm256_set1_epi8( match );
-    auto ptr = line;
-    for(;;)
-    {
-        __m256i l = _mm256_loadu_si256( (const __m256i*)ptr );
-        __m256i c = _mm256_cmpeq_epi8( l, m );
-        auto b = ~uint32_t( _mm256_movemask_epi8( c ) );
-        if( b != 0 )
-        {
-            line = ptr + __builtin_ctz( b );
-            return;
-        }
-        ptr += 32;
-    }
-}
-#else
-static inline void AdvanceToNot( const char*& line, char match )
-{
-    auto ptr = line;
-    for(;;)
-    {
-        uint64_t l;
-        memcpy( &l, ptr, 8 );
-        for( int i=0; i<8; i++ )
-        {
-            if( ( l & 0xFF ) != uint8_t( match ) )
-            {
-                line = ptr + i;
-                return;
-            }
-            l >>= 8;
-        }
-        ptr += 8;
-    }
-}
+                        uint64_t ip, t0;
+                        ring.Read( &ip, offset, sizeof( uint64_t ) );
+                        offset += sizeof( uint64_t );
+                        ring.Read( &t0, offset, sizeof( uint64_t ) );
+
+#if defined TRACY_HW_TIMER && ( defined __i386 || defined _M_IX86 || defined __x86_64__ || defined _M_X64 )
+                        t0 = ring.ConvertTimeToTsc( t0 );
 #endif
+                        QueueType type;
+                        switch( id )
+                        {
+                        case EventCpuCycles:
+                            type = QueueType::HwSampleCpuCycle;
+                            break;
+                        case EventInstructionsRetired:
+                            type = QueueType::HwSampleInstructionRetired;
+                            break;
+                        case EventCacheReference:
+                            type = QueueType::HwSampleCacheReference;
+                            break;
+                        case EventCacheMiss:
+                            type = QueueType::HwSampleCacheMiss;
+                            break;
+                        case EventBranchRetired:
+                            type = QueueType::HwSampleBranchRetired;
+                            break;
+                        case EventBranchMiss:
+                            type = QueueType::HwSampleBranchMiss;
+                            break;
+                        default:
+                            assert( false );
+                            break;
+                        }
 
-#ifdef __AVX2__
-template<int S>
-static inline void AdvanceTo( const char*& line, const char* match )
-{
-    auto first = uint8_t( match[0] );
-    __m256i m = _mm256_set1_epi8( first );
-    auto ptr = line;
-    for(;;)
-    {
-        __m256i l = _mm256_loadu_si256( (const __m256i*)ptr );
-        __m256i c = _mm256_cmpeq_epi8( l, m );
-        auto b = uint32_t( _mm256_movemask_epi8( c ) );
-        while( b != 0 )
-        {
-            auto bit = __builtin_ctz( b );
-            auto test = ptr + bit;
-            if( memcmp( test + 1, match + 1, S - 1 ) == 0 )
-            {
-                line = test;
-                return;
+                        TracyLfqPrepare( type );
+                        MemWrite( &item->hwSample.ip, ip );
+                        MemWrite( &item->hwSample.time, t0 );
+                        TracyLfqCommit;
+                    }
+                }
+                pos += hdr.size;
             }
-            b ^= ( 1u << bit );
+            assert( pos == end );
+            ring.Advance( end );
         }
-        ptr += 32;
-    }
-}
-#else
-template<int S>
-static inline void AdvanceTo( const char*& line, const char* match )
-{
-    auto first = uint8_t( match[0] );
-    auto ptr = line;
-    for(;;)
-    {
-        uint64_t l;
-        memcpy( &l, ptr, 8 );
-        for( int i=0; i<8; i++ )
+        if( !traceActive.load( std::memory_order_relaxed ) ) break;
+
+        if( s_ctxBufferIdx != s_numBuffers )
         {
-            if( ( l & 0xFF ) == first )
+            const auto ctxBufNum = s_numBuffers - s_ctxBufferIdx;
+
+            int activeNum = 0;
+            bool active[512];
+            uint32_t end[512];
+            uint32_t pos[512];
+            for( int i=0; i<ctxBufNum; i++ )
             {
-                if( memcmp( ptr + i + 1, match + 1, S - 1 ) == 0 )
+                const auto rbIdx = s_ctxBufferIdx + i;
+                const auto rbHead = s_ring[rbIdx].LoadHead();
+                const auto rbTail = s_ring[rbIdx].GetTail();
+                const auto rbActive = rbHead != rbTail;
+
+                active[i] = rbActive;
+                if( rbActive )
+                {
+                    activeNum++;
+                    end[i] = rbHead - rbTail;
+                    pos[i] = 0;
+                }
+                else
                 {
-                    line = ptr + i;
-                    return;
+                    end[i] = 0;
                 }
             }
-            l >>= 8;
-        }
-        ptr += 8;
-    }
-}
-#endif
-
-static void HandleTraceLine( const char* line )
-{
-    line += 23;
-    AdvanceTo( line, '[' );
-    line++;
-    const auto cpu = (uint8_t)ReadNumber( line );
-    line++;      // ']'
-    AdvanceToNot( line, ' ' );
+            if( activeNum > 0 )
+            {
+                hadData = true;
+                while( activeNum > 0 )
+                {
+                    int sel = -1;
+                    int64_t t0 = std::numeric_limits<int64_t>::max();
+                    for( int i=0; i<ctxBufNum; i++ )
+                    {
+                        if( !active[i] ) continue;
+                        auto rbPos = pos[i];
+                        assert( rbPos < end[i] );
+                        const auto rbIdx = s_ctxBufferIdx + i;
+                        perf_event_header hdr;
+                        s_ring[rbIdx].Read( &hdr, rbPos, sizeof( perf_event_header ) );
+                        if( hdr.type == PERF_RECORD_SAMPLE )
+                        {
+                            int64_t rbTime;
+                            s_ring[rbIdx].Read( &rbTime, rbPos + sizeof( perf_event_header ), sizeof( int64_t ) );
+                            if( rbTime < t0 )
+                            {
+                                t0 = rbTime;
+                                sel = i;
+                            }
+                        }
+                        else
+                        {
+                            rbPos += hdr.size;
+                            if( rbPos == end[i] )
+                            {
+                                active[i] = false;
+                                activeNum--;
+                            }
+                            else
+                            {
+                                pos[i] = rbPos;
+                            }
+                        }
+                    }
+                    assert( sel >= 0 || activeNum == 0 );
+                    if( sel >= 0 )
+                    {
+                        auto& ring = s_ring[s_ctxBufferIdx + sel];
+                        auto rbPos = pos[sel];
+                        auto offset = rbPos;
+                        perf_event_header hdr;
+                        ring.Read( &hdr, offset, sizeof( perf_event_header ) );
 
 #if defined TRACY_HW_TIMER && ( defined __i386 || defined _M_IX86 || defined __x86_64__ || defined _M_X64 )
-    const auto time = ReadNumber( line );
-#else
-    const auto ts = ReadNumber( line );
-    line++;      // '.'
-    const auto tus = ReadNumber( line );
-    const auto time = ts * 1000000000ll + tus * 1000ll;
+                        t0 = ring.ConvertTimeToTsc( t0 );
 #endif
 
-    line += 2;   // ': '
-    if( memcmp( line, "sched_switch", 12 ) == 0 )
-    {
-        line += 14;
-
-        AdvanceTo<8>( line, "prev_pid" );
-        line += 9;
-
-        const auto oldPid = uint32_t( ReadNumber( line ) );
-        line++;
-
-        AdvanceTo<10>( line, "prev_state" );
-        line += 11;
-
-        const auto oldState = (uint8_t)ReadState( *line );
-        line += 5;
-
-        AdvanceTo<8>( line, "next_pid" );
-        line += 9;
-
-        const auto newPid = uint32_t( ReadNumber( line ) );
-
-        uint8_t reason = 100;
-
-        TracyLfqPrepare( QueueType::ContextSwitch );
-        MemWrite( &item->contextSwitch.time, time );
-        MemWrite( &item->contextSwitch.oldThread, oldPid );
-        MemWrite( &item->contextSwitch.newThread, newPid );
-        MemWrite( &item->contextSwitch.cpu, cpu );
-        MemWrite( &item->contextSwitch.reason, reason );
-        MemWrite( &item->contextSwitch.state, oldState );
-        TracyLfqCommit;
-    }
-    else if( memcmp( line, "sched_wakeup", 12 ) == 0 )
-    {
-        line += 14;
-
-        AdvanceTo<4>( line, "pid=" );
-        line += 4;
-
-        const auto pid = uint32_t( ReadNumber( line ) );
-
-        TracyLfqPrepare( QueueType::ThreadWakeup );
-        MemWrite( &item->threadWakeup.time, time );
-        MemWrite( &item->threadWakeup.thread, pid );
-        TracyLfqCommit;
-    }
-}
-
-#ifdef __ANDROID__
-static void ProcessTraceLines( int fd )
-{
-    // Linux pipe buffer is 64KB, additional 1KB is for unfinished lines
-    char* buf = (char*)tracy_malloc( (64+1)*1024 );
-    char* line = buf;
-
-    for(;;)
-    {
-        if( !traceActive.load( std::memory_order_relaxed ) ) break;
+                        if( ring.GetId() == EventContextSwitch )
+                        {
+                            // Layout:
+                            //   u64 time
+                            //   u32 size
+                            //   u8  data[size]
+                            // Data (not ABI stable, but has not changed since it was added, in 2009):
+                            //   u8  hdr[8]
+                            //   u8  prev_comm[16]
+                            //   u32 prev_pid
+                            //   u32 prev_prio
+                            //   u64 prev_state
+                            //   u8  next_comm[16]
+                            //   u32 next_pid
+                            //   u32 next_prio
+
+                            offset += sizeof( perf_event_header ) + sizeof( uint64_t ) + sizeof( uint32_t ) + 8 + 16;
+
+                            uint32_t prev_pid, next_pid;
+                            uint64_t prev_state;
+
+                            ring.Read( &prev_pid, offset, sizeof( uint32_t ) );
+                            offset += sizeof( uint32_t ) + sizeof( uint32_t );
+                            ring.Read( &prev_state, offset, sizeof( uint64_t ) );
+                            offset += sizeof( uint64_t ) + 16;
+                            ring.Read( &next_pid, offset, sizeof( uint32_t ) );
+
+                            uint8_t reason = 100;
+                            uint8_t state;
+
+                            if(      prev_state & 0x0001 ) state = 104;
+                            else if( prev_state & 0x0002 ) state = 101;
+                            else if( prev_state & 0x0004 ) state = 105;
+                            else if( prev_state & 0x0008 ) state = 106;
+                            else if( prev_state & 0x0010 ) state = 108;
+                            else if( prev_state & 0x0020 ) state = 109;
+                            else if( prev_state & 0x0040 ) state = 110;
+                            else if( prev_state & 0x0080 ) state = 102;
+                            else                           state = 103;
+
+                            TracyLfqPrepare( QueueType::ContextSwitch );
+                            MemWrite( &item->contextSwitch.time, t0 );
+                            MemWrite( &item->contextSwitch.oldThread, prev_pid );
+                            MemWrite( &item->contextSwitch.newThread, next_pid );
+                            MemWrite( &item->contextSwitch.cpu, uint8_t( ring.GetCpu() ) );
+                            MemWrite( &item->contextSwitch.reason, reason );
+                            MemWrite( &item->contextSwitch.state, state );
+                            TracyLfqCommit;
+                        }
+                        else
+                        {
+                            assert( ring.GetId() == EventWakeup );
 
-        const auto rd = read( fd, line, 64*1024 );
-        if( rd <= 0 ) break;
+                            // Layout:
+                            //   u64 time
+                            //   u32 size
+                            //   u8  data[size]
+                            // Data:
+                            //   u8  hdr[8]
+                            //   u8  comm[16]
+                            //   u32 pid
+                            //   u32 prio
+                            //   u64 target_cpu
+
+                            offset += sizeof( perf_event_header ) + sizeof( uint64_t ) + sizeof( uint32_t ) + 8 + 16;
+
+                            uint32_t pid;
+                            ring.Read( &pid, offset, sizeof( uint32_t ) );
+
+                            TracyLfqPrepare( QueueType::ThreadWakeup );
+                            MemWrite( &item->threadWakeup.time, t0 );
+                            MemWrite( &item->threadWakeup.thread, pid );
+                            TracyLfqCommit;
+                        }
 
-#ifdef TRACY_ON_DEMAND
-        if( !GetProfiler().IsConnected() )
-        {
-            if( rd < 64*1024 )
-            {
-                assert( line[rd-1] == '\n' );
-                line = buf;
-                std::this_thread::sleep_for( std::chrono::milliseconds( 10 ) );
-            }
-            else
-            {
-                const auto end = line + rd;
-                line = end - 1;
-                while( line > buf && *line != '\n' ) line--;
-                if( line > buf )
+                        rbPos += hdr.size;
+                        if( rbPos == end[sel] )
+                        {
+                            active[sel] = false;
+                            activeNum--;
+                        }
+                        else
+                        {
+                            pos[sel] = rbPos;
+                        }
+                    }
+                }
+                for( int i=0; i<ctxBufNum; i++ )
                 {
-                    line++;
-                    const auto lsz = end - line;
-                    memmove( buf, line, lsz );
-                    line = buf + lsz;
+                    if( end[i] != 0 ) s_ring[s_ctxBufferIdx + i].Advance( end[i] );
                 }
             }
-            continue;
         }
-#endif
-
-        const auto end = line + rd;
-        line = buf;
-        for(;;)
-        {
-            auto next = (char*)memchr( line, '\n', end - line );
-            if( !next )
-            {
-                const auto lsz = end - line;
-                memmove( buf, line, lsz );
-                line = buf + lsz;
-                break;
-            }
-            HandleTraceLine( line );
-            line = ++next;
-        }
-        if( rd < 64*1024 )
-        {
-            std::this_thread::sleep_for( std::chrono::milliseconds( 10 ) );
-        }
-    }
-
-    tracy_free( buf );
-}
-
-void SysTraceWorker( void* ptr )
-{
-    ThreadExitHandler threadExitHandler;
-    SetThreadName( "Tracy SysTrace" );
-    int pipefd[2];
-    if( pipe( pipefd ) == 0 )
-    {
-        const auto pid = fork();
-        if( pid == 0 )
-        {
-            // child
-            close( pipefd[0] );
-            dup2( open( "/dev/null", O_WRONLY ), STDERR_FILENO );
-            if( dup2( pipefd[1], STDOUT_FILENO ) >= 0 )
-            {
-                close( pipefd[1] );
-                sched_param sp = { 4 };
-                pthread_setschedparam( pthread_self(), SCHED_FIFO, &sp );
-#if defined __ANDROID__ && ( defined __aarch64__ || defined __ARM_ARCH )
-                execlp( "su", "su", "root", "sh", "-c", "/data/tracy_systrace", (char*)nullptr );
-#endif
-                execlp( "su", "su", "root", "sh", "-c", "cat /sys/kernel/debug/tracing/trace_pipe", (char*)nullptr );
-                exit( 1 );
-            }
-        }
-        else if( pid > 0 )
-        {
-            // parent
-            close( pipefd[1] );
-            sched_param sp = { 5 };
-            pthread_setschedparam( pthread_self(), SCHED_FIFO, &sp );
-            ProcessTraceLines( pipefd[0] );
-            close( pipefd[0] );
-            waitpid( pid, nullptr, 0 );
-        }
-    }
-}
-#else
-static void ProcessTraceLines( int fd )
-{
-    // 32 bytes buffer space for wide unbound reads
-    char* buf = (char*)tracy_malloc( 64*1024 + 32 );
-
-    struct pollfd pfd;
-    pfd.fd = fd;
-    pfd.events = POLLIN | POLLERR;
-
-    for(;;)
-    {
-        while( poll( &pfd, 1, 0 ) <= 0 )
+        if( !traceActive.load( std::memory_order_relaxed ) ) break;
+        if( !hadData )
         {
-            if( !traceActive.load( std::memory_order_relaxed ) ) break;
             std::this_thread::sleep_for( std::chrono::milliseconds( 10 ) );
         }
-
-        const auto rd = read( fd, buf, 64*1024 );
-        if( rd <= 0 ) break;
-
-#ifdef TRACY_ON_DEMAND
-        if( !GetProfiler().IsConnected() ) continue;
-#endif
-
-        auto line = buf;
-        const auto end = buf + rd;
-        for(;;)
-        {
-            auto next = (char*)memchr( line, '\n', end - line );
-            if( !next ) break;
-            HandleTraceLine( line );
-            line = ++next;
-        }
     }
 
-    tracy_free( buf );
+    for( int i=0; i<s_numBuffers; i++ ) s_ring[i].~RingBuffer<RingBufSize>();
+    tracy_free_fast( s_ring );
 }
 
-void SysTraceWorker( void* ptr )
-{
-    ThreadExitHandler threadExitHandler;
-    SetThreadName( "Tracy SysTrace" );
-    char tmp[256];
-    memcpy( tmp, BasePath, sizeof( BasePath ) - 1 );
-    memcpy( tmp + sizeof( BasePath ) - 1, TracePipe, sizeof( TracePipe ) );
-
-    int fd = open( tmp, O_RDONLY );
-    if( fd < 0 ) return;
-    sched_param sp = { 5 };
-    pthread_setschedparam( pthread_self(), SCHED_FIFO, &sp );
-    ProcessTraceLines( fd );
-    close( fd );
-}
-#endif
-
 void SysTraceGetExternalName( uint64_t thread, const char*& threadName, const char*& name )
 {
     FILE* f;
diff --git a/client/TracySysTracePayload.hpp b/client/TracySysTracePayload.hpp
deleted file mode 100644
index 7c292f9d..00000000
--- a/client/TracySysTracePayload.hpp
+++ /dev/null
@@ -1,78 +0,0 @@
-// File: 'extra/systrace/tracy_systrace.armv7' (1149 bytes)
-// File: 'extra/systrace/tracy_systrace.aarch64' (1650 bytes)
-
-// Exported using binary_to_compressed_c.cpp
-
-namespace tracy
-{
-
-static const unsigned int tracy_systrace_armv7_size = 1149;
-static const unsigned int tracy_systrace_armv7_data[1152/4] =
-{
-    0x464c457f, 0x00010101, 0x00000000, 0x00000000, 0x00280003, 0x00000001, 0x000001f0, 0x00000034, 0x00000000, 0x05000200, 0x00200034, 0x00280007, 
-    0x00000000, 0x00000006, 0x00000034, 0x00000034, 0x00000034, 0x000000e0, 0x000000e0, 0x00000004, 0x00000004, 0x00000003, 0x00000114, 0x00000114, 
-    0x00000114, 0x00000013, 0x00000013, 0x00000004, 0x00000001, 0x00000001, 0x00000000, 0x00000000, 0x00000000, 0x000003fd, 0x000003fd, 0x00000005, 
-    0x00001000, 0x00000001, 0x000003fd, 0x000013fd, 0x000013fd, 0x00000080, 0x000000b3, 0x00000006, 0x00001000, 0x00000002, 0x00000400, 0x00001400, 
-    0x00001400, 0x0000007d, 0x000000b0, 0x00000006, 0x00000004, 0x6474e551, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000006, 
-    0x00000004, 0x70000001, 0x000003a4, 0x000003a4, 0x000003a4, 0x00000008, 0x00000008, 0x00000004, 0x00000004, 0x7379732f, 0x2f6d6574, 0x2f6e6962, 
-    0x6b6e696c, 0x00007265, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000001, 0x00000000, 0x00000000, 0x00000012, 0x00000016, 0x00000000, 
-    0x00000000, 0x00000012, 0x6f6c6400, 0x006e6570, 0x4342494c, 0x62696c00, 0x732e6c64, 0x6c64006f, 0x006d7973, 0x00000001, 0x00000003, 0x00000001, 
-    0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00010001, 0x0000000d, 0x00000010, 0x00000000, 0x00050d63, 0x00020000, 0x00000008, 
-    0x00000000, 0x000014bc, 0x00000116, 0x000014c0, 0x00000216, 0xe52de004, 0xe59fe004, 0xe08fe00e, 0xe5bef008, 0x000012dc, 0xe28fc600, 0xe28cca01, 
-    0xe5bcf2dc, 0xe28fc600, 0xe28cca01, 0xe5bcf2d4, 0xe92d4ff0, 0xe28db01c, 0xe24dd024, 0xe24dd801, 0xe59f017c, 0xe3a01001, 0xe3a08001, 0xe08f0000, 
-    0xebfffff0, 0xe59f116c, 0xe1a04000, 0xe08f1001, 0xebffffef, 0xe59f1160, 0xe1a06000, 0xe1a00004, 0xe08f1001, 0xebffffea, 0xe59f1150, 0xe1a07000, 
-    0xe1a00004, 0xe08f1001, 0xebffffe5, 0xe59f1140, 0xe1a05000, 0xe1a00004, 0xe08f1001, 0xebffffe0, 0xe58d0004, 0xe1a00004, 0xe59f1128, 0xe08f1001, 
-    0xebffffdb, 0xe59f1120, 0xe1a0a000, 0xe1a00004, 0xe08f1001, 0xebffffd6, 0xe1a04000, 0xe59f010c, 0xe3a01000, 0xe3a09000, 0xe08f0000, 0xe12fff36, 
-    0xe1a06000, 0xe3700001, 0xca000001, 0xe3a00000, 0xe12fff37, 0xe3a00009, 0xe3a01001, 0xe1cd01bc, 0xe3a00008, 0xe1cd01b4, 0xe3090680, 0xe3400098, 
-    0xe3a02000, 0xe58d000c, 0xe28d0010, 0xe58d7000, 0xe58d6018, 0xe58d8010, 0xe58d9008, 0xe12fff35, 0xe3500000, 0xca00001d, 0xe28d7018, 0xe28d8010, 
-    0xe28d9020, 0xe1a00007, 0xe3a01001, 0xe3a02000, 0xe12fff35, 0xe3500000, 0xda00000a, 0xe1a00006, 0xe1a01009, 0xe3a02801, 0xe12fff3a, 0xe3500001, 
-    0xba00000e, 0xe1a02000, 0xe3a00001, 0xe1a01009, 0xe12fff34, 0xea000003, 0xe59d2004, 0xe28d0008, 0xe3a01000, 0xe12fff32, 0xe1a00008, 0xe3a01001, 
-    0xe3a02000, 0xe12fff35, 0xe3500001, 0xbaffffe4, 0xe59d1000, 0xe3a00000, 0xe12fff31, 0xe24bd01c, 0xe8bd8ff0, 0x00000198, 0x00000190, 0x00000181, 
-    0x00000172, 0x00000163, 0x00000159, 0x0000014a, 0x00000138, 0x7ffffe4c, 0x00000001, 0x6362696c, 0x006f732e, 0x6e65706f, 0x69786500, 0x6f700074, 
-    0x6e006c6c, 0x736f6e61, 0x7065656c, 0x61657200, 0x72770064, 0x00657469, 0x7379732f, 0x72656b2f, 0x2f6c656e, 0x75626564, 0x72742f67, 0x6e696361, 
-    0x72742f67, 0x5f656361, 0x65706970, 0x00000000, 0x00000003, 0x000014b0, 0x00000002, 0x00000010, 0x00000017, 0x000001b4, 0x00000014, 0x00000011, 
-    0x00000015, 0x00000000, 0x00000006, 0x00000128, 0x0000000b, 0x00000010, 0x00000005, 0x00000158, 0x0000000a, 0x0000001c, 0x6ffffef5, 0x00000174, 
-    0x00000001, 0x0000000d, 0x0000001e, 0x00000008, 0x6ffffffb, 0x00000001, 0x6ffffff0, 0x0000018c, 0x6ffffffe, 0x00000194, 0x6fffffff, 0x00000001, 
-};
-
-static const unsigned int tracy_systrace_aarch64_size = 1650;
-static const unsigned int tracy_systrace_aarch64_data[1652/4] =
-{
-    0x464c457f, 0x00010102, 0x00000000, 0x00000000, 0x00b70003, 0x00000001, 0x000002e0, 0x00000000, 0x00000040, 0x00000000, 0x00000000, 0x00000000, 
-    0x00000000, 0x00380040, 0x00400006, 0x00000000, 0x00000006, 0x00000005, 0x00000040, 0x00000000, 0x00000040, 0x00000000, 0x00000040, 0x00000000, 
-    0x00000150, 0x00000000, 0x00000150, 0x00000000, 0x00000008, 0x00000000, 0x00000003, 0x00000004, 0x00000190, 0x00000000, 0x00000190, 0x00000000, 
-    0x00000190, 0x00000000, 0x00000015, 0x00000000, 0x00000015, 0x00000000, 0x00000001, 0x00000000, 0x00000001, 0x00000005, 0x00000000, 0x00000000, 
-    0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x000004e1, 0x00000000, 0x000004e1, 0x00000000, 0x00001000, 0x00000000, 0x00000001, 0x00000006, 
-    0x000004e8, 0x00000000, 0x000014e8, 0x00000000, 0x000014e8, 0x00000000, 0x0000018a, 0x00000000, 0x00000190, 0x00000000, 0x00001000, 0x00000000, 
-    0x00000002, 0x00000006, 0x000004e8, 0x00000000, 0x000014e8, 0x00000000, 0x000014e8, 0x00000000, 0x00000160, 0x00000000, 0x00000160, 0x00000000, 
-    0x00000008, 0x00000000, 0x6474e551, 0x00000006, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 
-    0x00000000, 0x00000000, 0x00000008, 0x00000000, 0x7379732f, 0x2f6d6574, 0x2f6e6962, 0x6b6e696c, 0x34367265, 0x00000000, 0x00000001, 0x00000001, 
-    0x00000001, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 
-    0x00000000, 0x00090003, 0x000002e0, 0x00000000, 0x00000000, 0x00000000, 0x00000010, 0x00000012, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 
-    0x0000000a, 0x00000012, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x62696c00, 0x732e6c64, 0x6c64006f, 0x006d7973, 0x706f6c64, 0x4c006e65, 
-    0x00434249, 0x00000000, 0x00000000, 0x00000000, 0x00010001, 0x00000001, 0x00000010, 0x00000000, 0x00050d63, 0x00020000, 0x00000017, 0x00000000, 
-    0x00001668, 0x00000000, 0x00000402, 0x00000002, 0x00000000, 0x00000000, 0x00001670, 0x00000000, 0x00000402, 0x00000003, 0x00000000, 0x00000000, 
-    0xa9bf7bf0, 0xb0000010, 0xf9433211, 0x91198210, 0xd61f0220, 0xd503201f, 0xd503201f, 0xd503201f, 0xb0000010, 0xf9433611, 0x9119a210, 0xd61f0220, 
-    0xb0000010, 0xf9433a11, 0x9119c210, 0xd61f0220, 0xa9bb67fc, 0xa9015ff8, 0xa90257f6, 0xa9034ff4, 0xa9047bfd, 0x910103fd, 0xd14043ff, 0xd10083ff, 
-    0x90000000, 0x91124000, 0x52800021, 0x52800039, 0x97ffffec, 0x90000001, 0x91126021, 0xaa0003f7, 0x97ffffec, 0x90000001, 0xaa0003f8, 0x91127421, 
-    0xaa1703e0, 0x97ffffe7, 0x90000001, 0xaa0003f3, 0x91128821, 0xaa1703e0, 0x97ffffe2, 0x90000001, 0xaa0003f4, 0x91129c21, 0xaa1703e0, 0x97ffffdd, 
-    0x90000001, 0xaa0003f5, 0x9112c421, 0xaa1703e0, 0x97ffffd8, 0x90000001, 0xaa0003f6, 0x9112d821, 0xaa1703e0, 0x97ffffd3, 0xaa0003f7, 0x90000000, 
-    0x9112f000, 0x2a1f03e1, 0xd63f0300, 0x2a0003f8, 0x36f80060, 0x2a1f03e0, 0xd63f0260, 0x90000009, 0x3dc12120, 0x52800128, 0x79003be8, 0x52800108, 
-    0x910043e0, 0x52800021, 0x2a1f03e2, 0xb9001bf8, 0xb90013f9, 0x79002be8, 0x3d8003e0, 0xd63f0280, 0x7100001f, 0x5400036c, 0x910063e0, 0x52800021, 
-    0x2a1f03e2, 0xd63f0280, 0x7100001f, 0x5400018d, 0x910083e1, 0x52a00022, 0x2a1803e0, 0xd63f02c0, 0xf100041f, 0x540001eb, 0xaa0003e2, 0x910083e1, 
-    0x52800020, 0xd63f02e0, 0x14000004, 0x910003e0, 0xaa1f03e1, 0xd63f02a0, 0x910043e0, 0x52800021, 0x2a1f03e2, 0xd63f0280, 0x7100041f, 0x54fffceb, 
-    0x2a1f03e0, 0xd63f0260, 0x914043ff, 0x910083ff, 0xa9447bfd, 0xa9434ff4, 0xa94257f6, 0xa9415ff8, 0xa8c567fc, 0xd65f03c0, 0x00000000, 0x00000000, 
-    0x00000000, 0x00000000, 0x00989680, 0x00000000, 0x6362696c, 0x006f732e, 0x6e65706f, 0x69786500, 0x6f700074, 0x6e006c6c, 0x736f6e61, 0x7065656c, 
-    0x61657200, 0x72770064, 0x00657469, 0x7379732f, 0x72656b2f, 0x2f6c656e, 0x75626564, 0x72742f67, 0x6e696361, 0x72742f67, 0x5f656361, 0x65706970, 
-    0x00000000, 0x00000000, 0x00000001, 0x00000000, 0x00000001, 0x00000000, 0x6ffffef5, 0x00000000, 0x000001a8, 0x00000000, 0x00000005, 0x00000000, 
-    0x00000228, 0x00000000, 0x00000006, 0x00000000, 0x000001c8, 0x00000000, 0x0000000a, 0x00000000, 0x0000001c, 0x00000000, 0x0000000b, 0x00000000, 
-    0x00000018, 0x00000000, 0x00000015, 0x00000000, 0x00000000, 0x00000000, 0x00000003, 0x00000000, 0x00001650, 0x00000000, 0x00000002, 0x00000000, 
-    0x00000030, 0x00000000, 0x00000014, 0x00000000, 0x00000007, 0x00000000, 0x00000017, 0x00000000, 0x00000270, 0x00000000, 0x0000001e, 0x00000000, 
-    0x00000008, 0x00000000, 0x6ffffffb, 0x00000000, 0x00000001, 0x00000000, 0x6ffffffe, 0x00000000, 0x00000250, 0x00000000, 0x6fffffff, 0x00000000, 
-    0x00000001, 0x00000000, 0x6ffffff0, 0x00000000, 0x00000244, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 
-    0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 
-    0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 0x00000000, 
-    0x00000000, 0x00000000, 0x000002a0, 0x00000000, 0x000002a0, 
-};
-
-}