Pinned host memory from a process-wide pool; the compressed chunk uploaded from pinned staging

The GPU engines are built per worker per step, and each took its page-locked host buffers with
cudaHostAlloc / cudaHostRegister and gave them back with cudaFreeHost / cudaHostUnregister. Those
calls map and lock pages under the process's memory-map lock, so all workers queue behind each
other: on the 5080 workstation a myoglobin run made ~3400 of them (6-22 thread-s, depending on the
load), 9gdj ~10400 (124 thread-s); on a four-socket host they cost 20-30 ms each.

- CudaHostPtr takes its memory from a pool: a freed buffer is kept in a power-of-two size class for
  the next engine, up to 512 MiB in all; a buffer above 64 MiB is never kept. Rugnux::Run hands the
  pool back at its end.
- PinnedVector (std::vector with the pool's allocator) replaces the registered vectors of
  AdaptiveSpotFinderGPU and ImagePreprocessorGPU; the Bragg predictors download into a pooled
  pinned buffer and copy into their output, instead of registering it.
- BSLZ4DecoderGPU copies the compressed chunk into a pooled pinned buffer before the upload, so it
  is one DMA instead of the driver's pageable staging (9gdj: 163 GB pageable H2D, 52 s of copy
  time, -> 147 GB pinned, 13 s).

Measured with nsys on the workstation: myob 3398 -> 1396 calls (674 of them the release at the
end), cytc 2986 -> 1270, 9gdj 10428 -> 1748 (124 -> 3.7 thread-s). Output identical (p.hkl md5 on
myob, cytc, 9gdj; CI tier battery).

Co-Authored-By: Claude Opus 5.5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01EizBKhTqrAYago9KJG3dA3

image_analysis/image_preprocessing/BSLZ4DecoderGPU.cu

@@ -2,6 +2,7 @@
 // SPDX-License-Identifier: GPL-3.0-only
 
 #include <algorithm>
+#include <cstring>
 #include <type_traits>
 
 #include "BSLZ4DecoderGPU.h"
@@ -146,6 +147,7 @@ void BSLZ4DecoderGPU::EnsureCompressedCapacity(size_t bytes) {
     const size_t want = bytes + bytes / 4;
     cuda_err(cudaStreamSynchronize(*stream));   // nothing may still be reading the old buffer
     gpu_compressed = CudaDevicePtr<uint8_t>(want);
+    host_compressed = CudaHostPtr<uint8_t>(want);
     compressed_capacity = want;
 }
 
@@ -238,7 +240,8 @@ BSLZ4ShuffledImage BSLZ4DecoderGPU::PrepareChunk(const CompressedImage &image) {
     host_status.get()[0] = 0;
     cuda_err(cudaMemcpyAsync(gpu_status.get(), host_status.get(), sizeof(uint32_t),
                              cudaMemcpyHostToDevice, *stream));
-    cuda_err(cudaMemcpyAsync(gpu_compressed.get(), src, clen, cudaMemcpyHostToDevice, *stream));
+    memcpy(host_compressed.get(), src, clen);
+    cuda_err(cudaMemcpyAsync(gpu_compressed.get(), host_compressed.get(), clen, cudaMemcpyHostToDevice, *stream));
 
     const int nb = static_cast<int>(nblk);
     if (nb > 0)