(alpha == 0.||K==0) return; int M4,N8=N&-8,K4; double *a_buffer = (double *)aligned_alloc(4096,K_BLOCKING*M_BLOCKING*sizeof(double)); double *b_buffer = (double *)aligned_alloc(4096,K_BLOCKING*N_BLOCKING*sizeof(double)); ING)?K_BLOCKING:K-k_count; m_inc=M>M_BLOCKING?M_BLOCKING:M; packing_a_k18(alpha,A+k_count*LDA,a_buffer,LDA,m_inc,k_inc); for (second_n_count=n_count;second_n_countbuffer+(second_n_count-n_count)*k_inc,LDB,k_inc,second_n_inc); macro_kernel_k18(a_buffer,b_buffer+(second_n_count-n_count)*k_inc,m_inc,second_n_inc,k_inc

_v, async_handler{}}; sycl::buffer bufA{dA.data(), sycl::range{dA.size()}}; sycl::buffer bufB{dB.data(), sycl::range{dB.size()}}; sycl::buffer bufO{dO.data(), sycl::range{dO.size()}}; _v, async_handler{}}; sycl::buffer bufA{dA.data(), sycl::range{dA.size()}}; sycl::buffer bufB{dB.data(), sycl::range{dB.size()}}; sycl::buffer bufO{dO.data(), sycl::range{dO.size()}}; _v, async_handler{}}; sycl::buffer bufA{dA.data(), sycl::range{dA.size()}}; sycl::buffer bufB{dB.data(), sycl::range{dB.size()}}; sycl::buffer bufO{dO.data(), sycl::range{dO.size()}};

8 } 9 10 void Main() 11 { 12 char buffer[16]{}; 13 14 // Prefer a range−based for−loop! 15 for(int i = 0; i < Size(buffer); ++i) { 16 // do something with buffer 17 } 18 } Andreas Fertig v1.0 B2B: fill buffer with data 13 } 14 15 void Main() 16 { 17 char buffer[1’024]{}; 18 19 Read(buffer, sizeof(buffer)); 20 Send(buffer, sizeof(buffer)); 21 22 char buffer2[2’048]{}; 23 24 Read(buffer, sizeof(buffer2)); sizeof(buffer2)); 25 Send(buffer, sizeof(buffer2)); 26 } Andreas Fertig v1.0 B2B: C++ Templates 17 Guidelines for efficient use of templates ­ An example ■ The pattern of passing value and length is:

Output • The output is a Data buffer. The buffer is a flat, contiguous block of data designed for storage or transmission. • As illustrated in the diagram, the data buffer is a single, uninterrupted space space used to hold data. Fanchen Su, XOffsetDatastructure, CppCon 2024 18 Data Buffer2.1.1 What Is Serialization and Deserialization • Process • The Process involves converting between structured First Category • Serialization • The output is a data buffer. • The process involves transforming the memory data structure into a data buffer. This transformation requires traversing and encoding

ProtocolHeader { std::array protocol_name; uint64_t magic_number; uint64_t buffer_size; uint32_t major_version; uint32_t minor_version; std::array queue_size_bytes; FastQueue – API struct QProducer { void Write(std::span buffer); }; struct QConsumer { int32_t TryRead(std::span buffer); // returns #bytes read, 0 if nothing to read };89 FastQueue Writer // simplified code! void QProducer::Write(std::span buffer) { const int32_t payloadSize = sizeof(int32_t) + buffer.size(); mLocalCounter += payloadSize; mQ->mWriteCounter.store(mLocalCounter

struct region { int size; }; void init_region(char* backing_buffer, size_t buffer_size ) { if(buffer_size < current_region.size) { LOG(“Buffer size too small”); return; } //other init code }An Introduction struct region { int size; }; void init_region(char* backing_buffer, size_t buffer_size ) { if(buffer_size < current_region.size) { LOG(“Buffer size too small”); return; } //other init code }An Introduction region { int size; }; void init_region(char* backing_buffer, size_t buffer_size ) { if(buffer_size < (size_t)current_region.size) { LOG(“Buffer size too small”); return; } //other init code }An Introduction

example● Unified shared memory provides an alternative pointer-based data management model to the buffer/accessor mode ○ Unified virtual address space (consistent pointers) ○ Pointer-based structures ○ ▪ CUDA & ROCm ▪ C _Complex ▪ Required reinterpret_cast from thrust::complex ▪ UB! 63BUFFER LIFETIME ▪sycl::buffer is a sometimes owning collection of data ▪sycl::accessor is a non-owning view that defines for the blocking behavior of a buffer destructor is that it blocks if there is some data to write back because a write accessor on it has been created, or if the buffer was constructed with attached host

control 0x04 (4) UCON control 0x08 (8) USTAT status 0x0C (12) UTXBUF transmit buffer 0x10 (16) URXBUF receive buffer 0x14 (20) UBRDIV baud rate divisor (control)What Volatile Means (and Doesn't Mean) character via the UART, you must access both the USTAT and UTXBUF registers.  The TBE bit (Transmit Buffer Empty) in USTAT is set to 1 when UTXBUF is ready for use.  You shouldn’t store a character into For example:  Writing to a control register may initiate an operation.  Reading from a receive buffer may set or clear bits in a status register.  Compiler optimizations might change the number of

switchingCallbacks vs coroutines // read length auto len = make_unique(); async_read(socket, buffer(len.get(), sizeof(int32_t)), [len = move(len)](auto ec, size_t len1){ if (!ec && len1 == sizeof(int32_t)) { // read data auto buf = make_unique(*len); async_read(socket, buffer(buf.get(), *len), [buf = move(buf), len = *len] (auto ec, size_t len2) { int32_t len; co_await async_read(s, buffer(&len, sizeof(len)), use_awaitable); // read data auto buf = make_unique(len); co_await async_read(s, buffer(buf.get(), len), use_awaitable);

CompletionToken> decltype(auto) async_read(AsyncReadStream& stream, DynamicBufferV2 buffer, CompletionToken&& token); 9 Dynamic Buffer Requirements struct /* ... */ { using const_buffers_type = /* ... */; Maximum of 4 294 967 295 (4 GiB) Naïve implementation could be induced to allocate extremely large buffer Denial of service attack if std::bad_alloc is thrown or OOM killer activates Must be able to limit ufferV2& buffer, std::size_t n, async_read_error err) { const auto max = buffer.max_size(); if ((n > max) || ((max - n) < buffer.size())) { return make_error_code(err); } buffer.grow(n);