leonarski_fandClaude Opus 5.5
c6f6396e46
A CUDA out-of-memory in an indexer worker was turned into an IndexerResult carrying only the message string. The rotation indexer then returned no lattice, the first pass treated that as a genuine negative and the rescues acted on it: in a stress run (two rugnux on one card) a run had both first-pass schemes fail with out-of-memory, the axis-sign rescue compared 0 frames at the flipped sign with the -1 "no lattice" sentinel, adopted the flipped sign, and the beam-centre rescue then moved the centre - a failure of the machine used as evidence about the crystal. The CUDA category was lost on the way, so the fail-fast check (IsFatalResourceError) never saw it. - IndexerResult::error is now the std::exception_ptr the worker caught. RotationIndexer::RunIndexing rethrows it before touching any state, so nothing is memoised for a failed run and the exception keeps its category: rugnux stops with the CUDA message at once (verified with a forced indexer failure: exit 1 after 1.9 s, "CUDA (GPU) error (...)"). FirstPass::indexer_error, RotationIndexer::GetIndexerError and RequireLattice's machine-failure branch are gone. - The beam-centre arm in RunAllPasses rethrows fatal resource errors like its sibling catches; it had caught the out-of-memory and carried on at the file's centre. - IndexAndRefine keeps per-frame indexer failures apart from frames that did not index (count + first exception). The frame stays "not attempted" as before; the receiver logs the count and the first error at the end of a collection (no cancel), rugnux's image pass and stills geometry refinement rethrow it and stop. A rotation-indexer failure inside IndexAndRefine (online rotation indexing, FinalizeRotationIndexing) is recorded the same way, so a live collection is not cancelled by it. - A failed device allocation or cuFFT plan now says how much was asked, how much of the card is free, and - when the card is merely full - that another program is using the GPU or the data set needs more than the card has, with the CPU remedy; previously "Failed to allocate device memory (out of memory)" / "Failed to create CUFFT plan with cufftPlanMany". Co-Authored-By: Claude Opus 5.5 (1M context) <noreply@anthropic.com>
88 lines
3.9 KiB
C++
88 lines
3.9 KiB
C++
// SPDX-FileCopyrightText: 2026 Filip Leonarski, Paul Scherrer Institute
|
|
// SPDX-License-Identifier: GPL-3.0-only
|
|
|
|
#include <catch2/catch_all.hpp>
|
|
#include "../common/CUDAWrapper.h"
|
|
|
|
#ifdef JFJOCH_USE_CUDA
|
|
|
|
#include <cuda_runtime.h>
|
|
#include <limits>
|
|
|
|
#include "../common/AzimuthalIntegrationMapping.h"
|
|
#include "../common/AzimuthalIntegrationProfile.h"
|
|
#include "../common/DiffractionExperiment.h"
|
|
#include "../common/PixelMask.h"
|
|
#include "../compression/JFJochCompressor.h"
|
|
#include "../image_analysis/IndexAndRefine.h"
|
|
#include "../image_analysis/MXAnalysisWithoutFPGA.h"
|
|
#include "../image_analysis/indexing/CUDAMemHelpers.h"
|
|
|
|
// What makes a handled CUDA failure dangerous: the failed call leaves the error behind as the
|
|
// thread's last error, and nothing but cudaGetLastError() takes it away - later successful calls
|
|
// neither clear it nor return it. So a route that failed, was caught and was replaced by another
|
|
// one hands its error to the next cudaGetLastError() in the code that followed, which is the check
|
|
// after a kernel launch. That is how a broker logged a raw "out of memory" over an image whose
|
|
// buffers were all allocated. cuda_clear_error() is what a handled failure discharges it with.
|
|
TEST_CASE("CudaDevicePtr_FailedAllocationLeavesErrorUntilCleared", "[CUDAMemHelpers]") {
|
|
if (get_gpu_count() == 0)
|
|
SKIP("No CUDA GPU present");
|
|
|
|
cuda_clear_error();
|
|
// Said as what it is, with the remedy, not as a bare CUDA status.
|
|
REQUIRE_THROWS_WITH(CudaDevicePtr<uint8_t>(std::numeric_limits<size_t>::max(), CudaAlloc::Synchronous),
|
|
Catch::Matchers::ContainsSubstring("another program is using this GPU"));
|
|
|
|
// Work that goes fine in between changes nothing: the error is still waiting.
|
|
CudaDevicePtr<uint8_t> fine(1 << 16);
|
|
REQUIRE(cudaStreamSynchronize(cuda_allocation_stream()) == cudaSuccess);
|
|
|
|
cuda_clear_error();
|
|
CHECK(cudaGetLastError() == cudaSuccess);
|
|
}
|
|
|
|
// The device decoder is allowed to fail - the host decompresses instead - but the host route runs
|
|
// kernels of its own and checks cudaGetLastError() after them. An image must come out of that
|
|
// whole, and with nothing left behind for the next one.
|
|
TEST_CASE("MXAnalysis_HandledDeviceDecodeFailureLeavesNoError", "[CUDAMemHelpers]") {
|
|
if (get_gpu_count() == 0)
|
|
SKIP("No CUDA GPU present");
|
|
|
|
constexpr size_t width = 1030, height = 1064;
|
|
|
|
DiffractionExperiment experiment(DetDECTRIS(width, height, "Test", {}));
|
|
experiment.ImagesPerTrigger(1).NumTriggers(1).PixelSigned(false)
|
|
.DetectorDistance_mm(75).BeamX_pxl(515).BeamY_pxl(532).IncidentEnergy_keV(12.4);
|
|
|
|
PixelMask pixel_mask(experiment);
|
|
AzimuthalIntegrationMapping mapping(experiment, pixel_mask);
|
|
IndexAndRefine index_and_refine(experiment, nullptr);
|
|
MXAnalysisWithoutFPGA analysis(experiment, mapping, pixel_mask, index_and_refine);
|
|
|
|
std::vector<uint16_t> pixels(width * height, 0);
|
|
for (size_t i = 0; i < pixels.size(); i += 997)
|
|
pixels[i] = 500;
|
|
JFJochBitShuffleCompressor compressor(CompressionAlgorithm::BSHUF_LZ4);
|
|
const std::vector<uint8_t> compressed = compressor.Compress(pixels);
|
|
REQUIRE(!compressed.empty());
|
|
|
|
DataMessage message{};
|
|
message.number = 0;
|
|
message.image = CompressedImage(compressed.data(), compressed.size(), width, height,
|
|
CompressedImageMode::Uint16, CompressionAlgorithm::BSHUF_LZ4);
|
|
|
|
SpotFindingSettings settings = DiffractionExperiment::DefaultDataProcessingSettings();
|
|
settings.enable = false;
|
|
|
|
// Stage the device route into failing, by leaving behind the error a request the card cannot
|
|
// serve leaves.
|
|
void *unservable = nullptr;
|
|
REQUIRE(cudaMalloc(&unservable, std::numeric_limits<size_t>::max()) != cudaSuccess);
|
|
|
|
AzimuthalIntegrationProfile profile(mapping);
|
|
REQUIRE_NOTHROW(analysis.Analyze(message, profile, settings));
|
|
CHECK(cudaGetLastError() == cudaSuccess);
|
|
}
|
|
|
|
#endif
|