B&B CPU determinism #798
B&B CPU determinism #798
Conversation
|
/ok to test 1b35ac9 |
|
/ok to test deff899 |
📝 WalkthroughWalkthroughAdds deterministic BSP branch-and-bound, shared work‑limit timers and scheduler, memory instrumentation (ins_vector), ML work‑unit predictors, extensive NVTX profiling, widespread dual‑simplex API refactors (templated vector types), MIP determinism/work‑limit plumbing, and deterministic reproducibility tests. Changes
Estimated code review effort🎯 5 (Critical) | ⏱️ ~120 minutes Possibly related PRs
Suggested reviewers
🚥 Pre-merge checks | ✅ 2 | ❌ 1❌ Failed checks (1 warning)
✅ Passed checks (2 passed)
✏️ Tip: You can configure your own custom pre-merge checks in the settings. ✨ Finishing touches
Thanks for using CodeRabbit! It's free for OSS, and your support helps us grow. If you like it, consider giving us a shout-out. Comment |
![coderabbitai[bot]](https://avatars.githubusercontent.com/in/347564?s=60&v=4){kind=small}
There was a problem hiding this comment.
Actionable comments posted: 20
Note
Due to the large number of review comments, Critical, Major severity comments were prioritized as inline comments.
Caution
Some comments are outside the diff and can’t be posted inline due to platform limitations.
⚠️ Outside diff range comments (3)
cpp/src/mip/local_search/local_search.cu (1)
57-58: Thread-safety concern: Global static variables.
local_search_best_objandpop_ptrare global static variables modified without synchronization. If multiple solver instances run concurrently (e.g., in a server context), this causes data races.Per coding guidelines: "Prevent thread-unsafe use of global and static variables; use proper mutex/synchronization in server code accessing shared solver state."
Consider making these instance members or using thread-local storage if per-thread isolation is acceptable.
cpp/src/mip/solve.cu (1)
198-231: Determinism flag is dropped after presolve.The reduced problem rebuild doesn’t pass the determinism flag, which can silently switch off deterministic mode post‑presolve.
🔧 Suggested fix
- problem = detail::problem_t<i_t, f_t>(result->reduced_problem); + problem = detail::problem_t<i_t, f_t>( + result->reduced_problem, + settings.get_tolerances(), + settings.determinism_mode == CUOPT_MODE_DETERMINISTIC);cpp/src/dual_simplex/branch_and_bound.cpp (1)
506-579:heuristic_preemption_callbackcan be invoked multiple times.
Line 524 calls it unconditionally and it is called again later when OPTIMAL/INFEASIBLE paths trigger. If the callback isn’t idempotent, this can double-stop heuristics.🧹 Suggested fix (call once)
- if (settings_.heuristic_preemption_callback != nullptr) { - settings_.heuristic_preemption_callback(); - } ... - if (settings_.heuristic_preemption_callback != nullptr) { - settings_.heuristic_preemption_callback(); - } ... - if (settings_.heuristic_preemption_callback != nullptr) { - settings_.heuristic_preemption_callback(); - } ... + if (settings_.heuristic_preemption_callback != nullptr) { + settings_.heuristic_preemption_callback(); + }
🤖 Fix all issues with AI agents
In `@cpp/src/dual_simplex/branch_and_bound.hpp`:
- Around line 374-380: The comparator node_ptr_lower_bound_comp currently only
compares mip_node_t::lower_bound causing non-deterministic ordering; update
operator() in node_ptr_lower_bound_comp to tie-break when lower_bound are equal
by comparing the BSP identity tuple fields (origin_worker_id then creation_seq)
from mip_node_t<i_t,f_t> so the ordering is deterministic and matches
backlog_node_compare_t / deterministic_node_ptr_comp, and ensure the heap_
instantiation continues to use node_ptr_lower_bound_comp.
In `@cpp/src/dual_simplex/pseudo_costs.cpp`:
- Around line 269-278: The new log unconditionally indexes solution[branch_var]
which is undefined when variable_selection_from_pseudo_costs returns -1 (e.g.,
when fractional is empty); update the code around branch_var (the call to
variable_selection_from_pseudo_costs and the log.printf) to check if branch_var
== -1 and handle that case (for example, log a message like "pc no branching
variable" or similar) instead of indexing solution, otherwise log the selected
variable and its value as before.
In `@cpp/src/dual_simplex/pseudo_costs.hpp`:
- Around line 35-63: In compute_pseudo_cost_averages,
num_initialized_down/num_initialized_up are incremented even when
pc_sum_down/pc_sum_up is non-finite; change the logic so you only increment
num_initialized_down when pc_num_down[j] > 0 AND std::isfinite(pc_sum_down[j])
(and likewise only increment num_initialized_up when pc_num_up[j] > 0 AND
std::isfinite(pc_sum_up[j])), leaving the fallback to f_t(1.0) unchanged when
the corresponding count remains zero; update the blocks around
pc_num_down/pc_sum_down and pc_num_up/pc_sum_up so the increment and the
accumulation happen together to avoid skewing averages.
In `@cpp/src/dual_simplex/simplex_solver_settings.hpp`:
- Around line 143-144: Initialize the uninitialized bool member `deterministic`
in the `SimplexSolverSettings` constructor initializer list (the same place
where `cudss_deterministic(false)` is set) to a sensible default (e.g., false)
so `deterministic` does not have an indeterminate value; update the constructor
for the `SimplexSolverSettings` class to include `deterministic(false)` in its
initializer list.
In `@cpp/src/dual_simplex/singletons.cpp`:
- Around line 203-211: The code unsafely uses static_cast<std::vector<i_t>&> on
col_perm and row_perm which is undefined when VectorI is an ins_vector wrapper;
replace those casts by calling the wrapper's .underlying() to get the actual
std::vector references (e.g., use col_perm.underlying() and
row_perm.underlying()) before passing into the row_col_graph_t<i_t> constructor
and any other places that currently cast the vectors (also update the second
occurrence around lines 237–246).
In `@cpp/src/mip/feasibility_jump/fj_cpu.cu`:
- Around line 517-521: The integrality check currently only treats
var_t::INTEGER as integer in is_integer_var(fj_cpu_climber_t<i_t, f_t>& fj_cpu,
i_t var_idx); update this function to also return true for binary variables by
checking fj_cpu.h_is_binary_variable[var_idx] in addition to
fj_cpu.h_var_types[var_idx] == var_t::INTEGER so binary variables (stored
separately with var_t::CONTINUOUS) are included in the integrality validation
used later in the integrality check logic.
In `@cpp/src/mip/utilities/cpu_worker_thread.cuh`:
- Around line 133-141: The method
cpu_worker_thread_base_t<Derived>::wait_for_cpu_solver uses std::chrono and the
CUOPT_LOG_DEBUG macro but the header does not include the required headers; add
the standard header <chrono> and the project's logging/header that defines
CUOPT_LOG_DEBUG (the logging header used elsewhere in the project) to the top of
cpu_worker_thread.cuh so the chrono types and logging macro are declared and
compilation won't depend on include order.
In `@cpp/src/utilities/hashing.hpp`:
- Around line 13-14: The compute_hash template currently takes its vector
parameter by value causing an unnecessary copy; change the signature of
compute_hash (template <typename i_t> inline uint32_t compute_hash(...)) to
accept the input as a const reference (const std::vector<i_t>& h_contents) and
update any internal usage accordingly to avoid mutating the vector and prevent
copies when hashing.
- Around line 28-33: compute_hash currently uses __CUDA_ARCH__ to conditionally
add __device__, which yields different signatures between host and device
builds; update the declaration of compute_hash (template<typename i_t> inline
uint32_t compute_hash(const i_t val) / function name compute_hash) to use a
CUDA-aware qualifier macro instead of __CUDA_ARCH__—e.g., wrap NVCC detection in
a compile-time macro and apply __host__ __device__ (or a custom CUDA_HOST_DEVICE
macro) unconditionally when compiling with NVCC so the signature is identical
for host and device; ensure the new macro is defined only when __CUDACC__ (or
equivalent) is present to avoid adding unknown attributes for non-CUDA
compilers.
In `@cpp/src/utilities/memory_instrumentation.hpp`:
- Line 26: Remove the leftover development comment "// Thank you Cursor!" from
the memory_instrumentation.hpp header—locate the stray comment in the file
(outside any function/class) and delete it so no debug/attribution comments
remain in the production code.
In `@cpp/src/utilities/models/fj_predictor/quantize.cpp`:
- Around line 1149-1161: The function fj_predictor::quantize lacks a bounds
check for the fid index, which can read past th_begin/th_len and cause UB; add a
guard at the top of quantize that validates fid is within the valid range (e.g.,
fid < th_begin.size() or fid < num_features) and returns the same sentinel/error
value (or handles error consistently) when out of range; ensure subsequent uses
of th_begin[fid], th_len[fid], and threshold[offset] are only done after this
check so no out-of-bounds access can occur.
In `@cpp/src/utilities/models/pdlp_predictor/quantize.cpp`:
- Around line 975-987: The quantize function lacks validation of the fid index
into th_begin/th_len/threshold and uses a magic sentinel 936; update
pdlp_predictor::quantize to first check that fid is within the valid range (0 <=
fid < th_begin.size() or th_len.size()) and handle out-of-range by returning the
sentinel or an error code, and replace the hard-coded 936 check with a derived
value (e.g., compare offset against total_num_threshold or threshold.size()) so
the "no thresholds for this feature" branch uses a boundary derived from the
actual arrays (use th_begin, th_len, threshold, or total_num_threshold
identifiers to locate and implement the checks).
In `@cpp/src/utilities/producer_sync.hpp`:
- Around line 19-52: The header is missing the <algorithm> include used by
std::find in producer_sync_t::deregister_producer; add `#include` <algorithm> at
the top of the file (alongside <vector>, <mutex>, etc.) so std::find is
available, then rebuild to ensure producer_sync_t::deregister_producer and any
other use of std::find compile on non-transitive STL implementations.
In `@cpp/src/utilities/timing_utils.hpp`:
- Around line 25-29: The rdtsc() inline uses GNU inline asm which breaks MSVC
x64; add an `#if/`#elif branch so rdtsc() uses the MSVC intrinsic __rdtsc()
(include <intrin.h>) when compiling under MSVC/_M_X64 and retain the inline asm
for GCC/Clang, and ensure headers are conditionally included. Also replace
non-portable printf("%lu", some_uint64) usage by printing the 64-bit value
portably—either use PRIx/PRIu macros from <inttypes.h> (e.g., PRIu64) or cast
the value to (unsigned long long) and use "%llu"—and update includes
accordingly.
In `@cpp/src/utilities/work_limit_timer.hpp`:
- Around line 29-43: Change global_work_units_elapsed to an atomic counter and
use relaxed atomics wherever threads read/write it: replace double
global_work_units_elapsed with std::atomic<double> (or an atomic integer with
fixed scaling if preferred), include <atomic>, update
work_limit_context_t::record_work to use fetch_add(work,
std::memory_order_relaxed) and then load the new value with
load(std::memory_order_relaxed) before calling
scheduler->on_work_recorded(*this, ...), and replace any direct
assignments/reads to work_limit_context_t::global_work_units_elapsed (e.g., the
write in work_unit_scheduler.cpp and all worker-thread reads in
branch_and_bound.cpp) with atomic store/load using std::memory_order_relaxed so
all accesses are synchronized without heavy ordering.
In `@cpp/src/utilities/work_unit_predictor.cpp`:
- Around line 44-52: The code writes different union members of model_t::Entry
(data[]) depending on presence of a feature then later reads the other member,
causing undefined behavior; change model_t::Entry to include an explicit
tag/flag (e.g., bool is_missing or an enum) or replace the union with a struct
containing std::optional<float> so you always have a well-defined way to test
presence, set that flag/optional when populating data[i] (use the iterator
returned by features.find to avoid calling features.at twice) and update the
later check (the place currently checking data[i].missing != -1) to use the new
flag/optional instead of reading a possibly-uninitialized union member; ensure
all code paths that construct or read model_t::Entry use the new presence
indicator.
- Around line 60-63: The code uses a 32-bit key from
cuopt::linear_programming::detail::compute_hash(cache_vec) to index
prediction_cache, which allows hash collisions to return wrong predictions;
change the cache to store and verify the full feature vector with each cached
value (e.g., map key -> pair<vector<float>, float> or store vector in the mapped
value) and on lookup (when computing key and finding cached_it) compare
cached_it->second's stored vector to cache_vec and only return the cached
prediction on exact match, otherwise treat as a miss and compute/store a new
entry; update any insertion logic to save the full cache_vec alongside the
prediction and adjust prediction_cache's type accordingly (alternatively change
the container to use the feature vector itself as the key).
In `@cpp/src/utilities/work_unit_predictor.hpp`:
- Around line 29-44: The constructor for cpu_work_unit_scaler_t divides
baseline_max_clock by get_cpu_max_clock_mhz() without validating the result,
which can produce inf/NaN if max_clock <= 0 or non-finite; update the
constructor to validate max_clock (call get_cpu_max_clock_mhz()), check for <= 0
or non-finite (use std::isfinite or equivalent) and if invalid set
scaling_factor_ to a safe default (e.g., 1.0) or compute using a clamped minimum
value, otherwise compute scaling_factor_ = baseline_max_clock / max_clock;
ensure scaling_factor_ is always initialized to a finite value.
- Around line 50-60: predict_scalar in work_unit_predictor_t mutates the mutable
prediction_cache without synchronization, causing data races when shared
predictors (e.g., static bs_predictor used by solve_node_bsp run via
run_worker_loop and predictors stored in context.work_unit_predictors) are
accessed by multiple worker threads; fix by making prediction_cache access
thread-safe—either add a mutable std::mutex (e.g., mutex_ in
work_unit_predictor_t) and lock it around reads/writes to prediction_cache
inside predict_scalar, or make the predictor/cache thread_local and ensure
bs_predictor and entries in context.work_unit_predictors are created per-thread;
update work_unit_predictor_t (prediction_cache and predict_scalar) and adjust
call sites (bs_predictor, context.work_unit_predictors) accordingly.
In `@cpp/src/utilities/work_unit_scheduler.cpp`:
- Around line 103-107: current_sync_target() and on_work_recorded() read
barrier_generation_ without synchronization while wait_at_sync_point() mutates
it under a mutex, causing a data race; fix by either acquiring the same mutex
(the mutex used in wait_at_sync_point()) around reads in current_sync_target()
and where on_work_recorded() logs barrier_generation_, or convert
barrier_generation_ to std::atomic<size_t> and use atomic reads; pick one
approach and update all read sites (current_sync_target(), on_work_recorded())
to use the chosen synchronization so reads are safe against concurrent writers.
🟡 Minor comments (15)
cpp/src/utilities/timing_utils.hpp-55-56 (1)
55-56: Use PRIu64 for portable 64-bit printf formatting.Lines 55-56 use
%luforuint64_tvariables (avg_##nameandmedian_##name), which is not portable. On Windows (LLP64 model),unsigned longis 32-bit whileuint64_tis 64-bit, causing truncation. UsePRIu64from<cinttypes>for portable formatting across all platforms.🔧 Proposed fix (portable formatting)
`#include` <cstdint> +#include <cinttypes> `#include` <cstdio> `#define` CYCLE_TIMING_EPILOGUE(name) \ do { \ uint64_t t_end_##name = cuopt::rdtsc(); \ timing_buffer_##name[timing_idx_##name++] = t_end_##name - t_start_##name; \ if (timing_idx_##name == timing_buffer_size_##name) { \ uint64_t sum_##name = 0; \ for (size_t i = 0; i < timing_buffer_size_##name; ++i) \ sum_##name += timing_buffer_##name[i]; \ uint64_t avg_##name = sum_##name / timing_buffer_size_##name; \ std::array<uint64_t, timing_buffer_size_##name> sorted_##name = timing_buffer_##name; \ std::nth_element(sorted_##name.begin(), \ sorted_##name.begin() + timing_buffer_size_##name / 2, \ sorted_##name.end()); \ uint64_t median_##name = sorted_##name[timing_buffer_size_##name / 2]; \ - printf(`#name` ": avg=%lu cycles, median=%lu cycles (n=%zu)\n", + printf(`#name` ": avg=%" PRIu64 " cycles, median=%" PRIu64 " cycles (n=%zu)\n", avg_##name, median_##name, timing_buffer_size_##name); \ timing_idx_##name = 0; \ } \ } while (0)cpp/src/mip/problem/problem_helpers.cuh-359-362 (1)
359-362: Preserve CSR validation for empty matrices.
The early return skipscheck_csr_representation, which can hide malformed offsets/indices. Consider keeping the validation even when skipping cuSPARSE.🧩 Proposed fix
- // skip if the matrix is empty - if (values.size() == 0) { return; } + // skip if the matrix is empty, but still validate CSR layout + if (values.size() == 0) { + check_csr_representation(values, offsets, indices, handle_ptr, cols, rows); + return; + }cpp/src/utilities/hashing.hpp-8-9 (1)
8-9: Missing<cstring>include forstd::memcpy.The code uses
std::memcpyon lines 20 and 37, but<cstring>is not included. This may compile on some toolchains due to transitive includes but is not portable.Proposed fix
`#include` <cstdint> +#include <cstring> `#include` <vector>cpp/src/utilities/models/dualsimplex_predictor/quantize.cpp-1860-1872 (1)
1860-1872: Missing bounds check onfidparameter.The
fidparameter indexes intoth_begin[]andth_len[]arrays (18 elements each, indices 0-17). If an invalidfid >= 18is passed, this causes undefined behavior via out-of-bounds array access.Consider adding a bounds check or documenting the valid range:
Proposed fix
int dualsimplex_predictor::quantize(float val, unsigned fid) { + // fid must be in range [0, 17] + if (fid >= 18) { return -10; } const size_t offset = th_begin[fid];cpp/src/utilities/models/cpufj_predictor/quantize.cpp-262-275 (1)
262-275: Add a defensivefidbounds check before indexing.
Line 264 indexesth_begin[fid]without validation; an invalidfidwould cause OOB reads. A cheap guard prevents UB if inputs drift.🐛 Proposed fix
int cpufj_predictor::quantize(double val, unsigned fid) { + if (fid >= NUM_FEATURES) { return -10; } const size_t offset = th_begin[fid]; const double* array = &threshold[offset]; int len = th_len[fid];cpp/src/utilities/memory_instrumentation.hpp-837-844 (1)
837-844: Disabled wrapper assumes methods exist that enabled wrapper guards with SFINAE.The disabled wrapper unconditionally provides
front(),back(),data(),max_size(),capacity(), etc., while the enabled wrapper uses SFINAE (has_front,has_data, etc.) to conditionally enable these methods. IfTlacks these methods, the disabled build will fail to compile while the enabled build succeeds.For consistency, the disabled wrapper should either:
- Also use SFINAE guards, or
- Document that
Tmust provide the fullstd::vector-like interfacecpp/src/utilities/memory_instrumentation.hpp-38-38 (1)
38-38: Instrumentation is hardcoded to enabled; should be configurable.
CUOPT_ENABLE_MEMORY_INSTRUMENTATIONis hardcoded to1. For production builds, this should likely be0by default or controlled via a CMake option to avoid any potential overhead. The documentation at lines 15-16 suggests it should be definable externally.Proposed fix
-#define CUOPT_ENABLE_MEMORY_INSTRUMENTATION 1 +#ifndef CUOPT_ENABLE_MEMORY_INSTRUMENTATION +#define CUOPT_ENABLE_MEMORY_INSTRUMENTATION 0 +#endifcpp/src/dual_simplex/basis_updates.cpp-21-22 (1)
21-22: Remove unusedins_vectoralias.The
ins_vectorimport on line 22 is not used anywhere in this file. Remove the unusedusingstatement.cpp/tests/linear_programming/c_api_tests/c_api_test.c-1256-1258 (1)
1256-1258: Use a fixed/explicit seed for the deterministic B&B test.The test explicitly validates deterministic behavior (
CUOPT_MODE_DETERMINISTICis set), but usesrand()without seeding. Sincesrand()is not called, the default seed is implementation-defined, causing the seed value to vary across platforms and runs. This undermines the test's ability to validate determinism consistently. Use a fixed constant instead.Proposed fix
- int seed = rand(); + const int seed = 12345; printf("Seed: %d\n", seed);cpp/src/mip/problem/problem.cu-881-928 (1)
881-928: Potential division by zero inunbalancednesscalculation.On line 926,
unbalancedness = nnz_stddev / mean;could produce infinity or NaN ifmeanis zero (which happens whensumis zero, i.e., an empty or all-zero-nnz-per-row matrix).While
n_constraints > 0is checked,sumcould still be zero if all rows have zero non-zeros (though this may be impossible by construction).🐛 Suggested defensive fix
nnz_stddev = std::sqrt(variance); - unbalancedness = nnz_stddev / mean; + unbalancedness = (mean > 0.0) ? (nnz_stddev / mean) : 0.0; }cpp/tests/mip/determinism_test.cu-158-174 (1)
158-174: Add test isolation forseed_generatorglobal state.The
seed_generatorclass uses a static member variable (seed_) that persists across test invocations. While this test resets the seed before each internal solve call, there is no test-level cleanup viaSetUp()orTearDown(). If other tests or test instances modify the seed, or if tests run in parallel, this could cause state leakage.For consistency with other tests in this file (e.g.,
reproducible_objectiveat line 56 usessettings.seed), consider either:
- Using
settings.seed = seed;instead of directseed_generator::set_seed()calls- Adding a
SetUp()method to initialize seed_generator state before each parameterized test instancecpp/src/dual_simplex/bb_event.hpp-142-153 (1)
142-153: Initialize horizon fields to avoid undefined values.
horizon_start/horizon_endare uninitialized unlessclear()is called first. Default initializers keep the struct safe in all construction paths.🔧 Proposed fix
struct bb_event_batch_t { std::vector<bb_event_t<i_t, f_t>> events; - double horizon_start; - double horizon_end; + double horizon_start = 0.0; + double horizon_end = 0.0;cpp/src/dual_simplex/vector_math.hpp-60-95 (1)
60-95: Add permutation index bounds validation to catch misuse early.The permutation helper functions access vectors using indices from the permutation array without bounds validation. While the internal LU and basis algorithms should generate valid permutations by construction, defensive bounds checks in these low-level utilities would catch index corruption or misuse deterministically. This aligns with the codebase's need for rigorous index validation across problem transformations (as evidenced by similar indexing issues found elsewhere in the dual_simplex module).
🔧 Proposed guardrails
template <typename VectorI, typename VectorF_in, typename VectorF_out> int permute_vector(const VectorI& p, const VectorF_in& b, VectorF_out& x) { auto n = p.size(); assert(x.size() == n); assert(b.size() == n); for (decltype(n) k = 0; k < n; ++k) { - x[k] = b[p[k]]; + auto idx = static_cast<decltype(n)>(p[k]); + assert(idx < n); + x[k] = b[idx]; } return 0; } @@ int inverse_permute_vector(const VectorI& p, const VectorF_in& b, VectorF_out& x) { auto n = p.size(); assert(x.size() == n); assert(b.size() == n); for (decltype(n) k = 0; k < n; ++k) { - x[p[k]] = b[k]; + auto idx = static_cast<decltype(n)>(p[k]); + assert(idx < n); + x[idx] = b[k]; } return 0; } @@ int inverse_permutation(const VectorI_in& p, VectorI_out& pinv) { auto n = p.size(); if (pinv.size() != n) { pinv.resize(n); } for (decltype(n) k = 0; k < n; ++k) { - pinv[p[k]] = k; + auto idx = static_cast<decltype(n)>(p[k]); + assert(idx < n); + pinv[idx] = k; } return 0; }cpp/src/mip/feasibility_jump/fj_cpu.cu-284-341 (1)
284-341: Guard feature stats against empty problems.
Lines 297–340 and 355–389 divide byn_vars/n_cstrs. If either is 0, this yields divide-by-zero and NaNs.🧯 Suggested guard
i_t total_nnz = fj_cpu.h_reverse_offsets.back(); i_t n_vars = fj_cpu.h_reverse_offsets.size() - 1; i_t n_cstrs = fj_cpu.h_offsets.size() - 1; + + if (n_vars == 0 || n_cstrs == 0) { + fj_cpu.avg_var_degree = 0.0; + fj_cpu.avg_cstr_degree = 0.0; + fj_cpu.max_var_degree = 0; + fj_cpu.max_cstr_degree = 0; + fj_cpu.var_degree_cv = 0.0; + fj_cpu.cstr_degree_cv = 0.0; + fj_cpu.problem_density = 0.0; + return; + }Also applies to: 350-389
benchmarks/linear_programming/cuopt/run_mip.cpp-202-212 (1)
202-212: Seed is forced even when determinism is off.
Line 211 setssettings.seed = 42unconditionally, which changes opportunistic runs. Consider setting the seed only whendeterministicis true (or expose a CLI seed).💡 Suggested change
- settings.seed = 42; + if (deterministic) { + settings.seed = 42; + }
🧹 Nitpick comments (19)
cpp/src/utilities/version_info.cpp (1)
166-201: Consider adding exception handling aroundstd::stodto prevent unexpected exceptions.The
std::stodcall on line 191 can throwstd::invalid_argumentorstd::out_of_rangeif/proc/cpuinfocontains malformed data. Since this occurs within athread_local staticlambda initialization, an uncaught exception could propagate unexpectedly. As per coding guidelines, prefer error handling over exceptions in utility paths.♻️ Suggested fix
if (line.find("cpu MHz") != std::string::npos) { std::size_t colon = line.find(':'); if (colon != std::string::npos) { - double mhz = std::stod(line.substr(colon + 1)); - if (mhz > max_mhz) { max_mhz = mhz; } + try { + double mhz = std::stod(line.substr(colon + 1)); + if (mhz > max_mhz) { max_mhz = mhz; } + } catch (...) { + // Skip malformed entries + } } }cpp/src/dual_simplex/mip_node.hpp (1)
280-291: Fix hash return type mismatch.
compute_path_hash()declares return typeuint64_t(line 280) but returnsdetail::compute_hash(path_steps)which isuint32_t(perutilities/hashing.hppline 14). The implicit conversion masks that only 32 bits of actual hash entropy exist, misleading about collision likelihood. Change return type touint32_tto match the implementation, or upgrade to a 64-bit hash if 64-bit collision resistance is required.cpp/src/utilities/models/fj_predictor/header.h (1)
20-25: Consider using C++ standard headers instead of C headers.For consistency with C++ best practices, consider using the C++ equivalents:
<cfloat>instead of<float.h><cmath>instead of<math.h><cstdint>instead of<stdint.h><cstdio>instead of<stdio.h><cstdlib>instead of<stdlib.h><cstring>instead of<string.h>This ensures names are in the
stdnamespace and is idiomatic C++.cpp/src/utilities/models/fj_predictor/quantize.cpp (1)
1162-1180: Consider extracting common quantize logic to a shared template.The binary search quantization logic is duplicated across
fj_predictor,pdlp_predictor,dualsimplex_predictor, and likely other predictors. Since only the data type and threshold arrays differ, a shared templated implementation could reduce maintenance burden.However, if these files are auto-generated from LightGBM model exports, keeping them self-contained may be preferable for regeneration simplicity. Based on learnings, code duplication in solver components (3+ occurrences) should be refactored into shared utilities.
cpp/src/utilities/memory_instrumentation.hpp (1)
406-416: Redundant preprocessor check inside already-conditional code block.The
#ifdef CUOPT_ENABLE_MEMORY_INSTRUMENTATIONcheck at line 409 is inside a class definition that's already inside#if CUOPT_ENABLE_MEMORY_INSTRUMENTATION(line 269). This inner check is always true when this code compiles.Proposed fix
auto operator*() const { if constexpr (IsConst) { -#ifdef CUOPT_ENABLE_MEMORY_INSTRUMENTATION wrapper_->byte_loads += sizeof(value_type); -#endif return *iter_; } else { return element_proxy_t(*iter_, *wrapper_); } }cpp/include/cuopt/linear_programming/mip/solver_solution.hpp (1)
26-34: Document the newWorkLimittermination status in public API docs.
This is a public header change; please update the relevant Doxygen/API docs or release notes so users discover the new status. As per coding guidelines, please update the public API documentation.cpp/include/cuopt/linear_programming/mip/solver_settings.hpp (1)
82-82: Consider adding documentation forwork_limit.For consistency with the documented
determinism_modeandseedfields, consider adding a Doxygen comment forwork_limitexplaining its purpose (resource budget measured in deterministic work units) and behavior (infinity = no limit).As per coding guidelines for public headers, new public functions/fields should have documentation comments.
📝 Suggested documentation
+ /** + * `@brief` Work limit for the MIP solver (deterministic work units). + * + * Limits solver execution based on deterministic work units rather than + * wall-clock time. Use infinity (default) for no limit. + */ f_t work_limit = std::numeric_limits<f_t>::infinity();cpp/src/mip/diversity/diversity_manager.cu (2)
309-335: Verify CPUFJ cleanup on early termination paths.The deterministic mode branch starts CPUFJ via
start_cpufj_deterministicand stops it viastop_cpufj_deterministic. However, if an exception is thrown between lines 323 and 331, or ifcheck_b_b_preemption()throws, the CPUFJ thread may not be properly stopped.Consider using an RAII guard (similar to
ls_cpufj_raii_guard_tused later in the non-deterministic path at line 375) to ensurestop_cpufj_deterministic()is called even on exceptional exit paths.♻️ Suggested RAII pattern
// Example RAII guard for deterministic CPUFJ struct deterministic_cpufj_guard_t { deterministic_cpufj_guard_t(local_search_t<i_t, f_t>& ls) : ls(ls) {} ~deterministic_cpufj_guard_t() { ls.stop_cpufj_deterministic(); } local_search_t<i_t, f_t>& ls; };
336-345: Environment variable bypass lacks RAII cleanup.The
CUOPT_DISABLE_GPU_HEURISTICSbranch at lines 336-345 spins incheck_b_b_preemption()but doesn't have any CPUFJ threads to clean up. The logic appears correct for debugging/testing B&B tree determinism in isolation.However, consider logging the bypass reason at DEBUG level rather than INFO to reduce log noise in production scenarios.
cpp/src/dual_simplex/solve.cpp (1)
193-205: NVTX scope label appears to be misnamed.The NVTX range is named "DualSimplex::phase2" at line 195, but this code block executes Phase 1 (dual_phase2 with argument
1indicating phase 1, per the function signature). Consider renaming to "DualSimplex::phase1" for clarity and accurate profiling attribution.♻️ Suggested fix
- raft::common::nvtx::range scope_phase1("DualSimplex::phase2"); + raft::common::nvtx::range scope_phase1("DualSimplex::phase1");cpp/tests/mip/determinism_test.cu (2)
55-86: Consider adding seed toreproducible_solution_vectortest for consistency.The
reproducible_objectiveandreproducible_high_contentiontests setsettings.seed, but thereproducible_solution_vectortest (lines 124-142) does not. For consistency and to ensure reproducibility testing is comprehensive, consider adding the seed parameter.Also, there's a stray semicolon on line 102.
♻️ Suggested fix
TEST_F(DeterministicBBTest, reproducible_solution_vector) { auto path = make_path_absolute("/mip/swath1.mps"); auto problem = mps_parser::parse_mps<int, double>(path, false); handle_.sync_stream(); mip_solver_settings_t<int, double> settings; settings.time_limit = 60.0; settings.determinism_mode = CUOPT_MODE_DETERMINISTIC; settings.num_cpu_threads = 8; settings.work_limit = 2; + + auto seed = std::random_device{}() & 0x7fffffff; + std::cout << "Tested with seed " << seed << "\n"; + settings.seed = seed; auto solution1 = solve_mip(&handle_, problem, settings);
100-104: Minor: Remove stray semicolon.Line 102 has an extra semicolon after the seed initialization.
auto seed = std::random_device{}() & 0x7fffffff; - ; std::cout << "Tested with seed " << seed << "\n";cpp/src/mip/solver_context.cuh (1)
62-68: Consider documenting the 5.0 second synchronization interval.The
work_unit_scheduler_is initialized with a 5.0 second default interval. This magic number could benefit from a named constant or comment explaining the rationale for this specific value.- // synchronization every 5 seconds for deterministic mode - work_unit_scheduler_t work_unit_scheduler_{5.0}; + // Synchronization interval for deterministic BSP mode. + // 5 seconds balances synchronization overhead vs. work distribution accuracy. + static constexpr double kDefaultSyncIntervalSeconds = 5.0; + work_unit_scheduler_t work_unit_scheduler_{kDefaultSyncIntervalSeconds};cpp/src/utilities/models/bounds_strengthening_predictor/header.h (1)
8-13: Consider using C++ standard headers instead of C headers.The file uses C-style headers (
<float.h>,<math.h>, etc.) instead of their C++ equivalents (<cfloat>,<cmath>,<cstdint>,<cstdio>,<cstdlib>,<cstring>). While functionally equivalent, C++ headers place symbols in thestdnamespace and are preferred in C++ codebases.However, if this header is auto-generated by LightGBM's model export tooling, modifying it may not be practical.
cpp/src/mip/local_search/local_search.cu (1)
151-191: Verifybbreference lifetime in callback.The improvement callback on line 182 captures
&bbby reference. The callback is stored indeterministic_cpu_fj.fj_cpu->improvement_callbackand may be invoked afterstart_cpufj_deterministicreturns.Ensure that
bboutlives the CPUFJ thread execution. Based on the code flow insolver.cu,branch_and_boundis kept alive untilbranch_and_bound_status_future.get()completes, which should be after CPUFJ termination. However, this implicit lifetime dependency could benefit from a comment.// Set up callback to send solutions to B&B with work unit timestamps + // Note: bb reference must remain valid until stop_cpufj_deterministic() is called deterministic_cpu_fj.fj_cpu->improvement_callback = [&bb](f_t obj, const std::vector<f_t>& h_vec, double work_units) {cpp/src/dual_simplex/branch_and_bound.hpp (1)
354-419: BSP state members are well-organized.The BSP-related private members properly separate concerns: worker pools, scheduler, synchronization primitives, heaps, and debug logging. The
queued_heuristic_solution_tstruct with work-unit timestamp enables deterministic solution ordering.Note: There are two
private:labels (line 144 and 354). While valid C++, this is unconventional. Consider whether the secondprivate:section is intentional for code organization or should be removed.cpp/src/mip/feasibility_jump/fj_cpu.cuh (1)
156-158: Consider documenting the work_unit_bias semantics.The comment says "< 1.0 means faster" but this could be clearer. A value of 0.8 means work is counted at 80% of its actual cost, effectively making CPUFJ appear to progress slower in work-unit terms, which keeps it "ahead" in real time relative to the B&B horizon.
cpp/src/dual_simplex/bsp_debug.hpp (1)
218-243: Consider logging a warning when directory creation fails.The
set_settingsmethod silently disables all features whencreate_directoriesthrows. This could make debugging difficult if users don't realize their debug output is being suppressed due to permission issues.Suggested enhancement
try { std::filesystem::create_directories(settings_.output_dir); } catch (const std::filesystem::filesystem_error& e) { + // Consider logging to stderr so users know debug output is disabled + std::fprintf(stderr, "BSP debug: failed to create directory '%s': %s\n", + settings_.output_dir.c_str(), e.what()); settings_.enable_event_log = false; settings_.enable_timeline = false; settings_.enable_tree_dot = false; settings_.enable_state_json = false; settings_.enable_determinism_trace = false; }cpp/src/dual_simplex/phase2.cpp (1)
30-39: Gate PHASE2_NVTX_RANGES behind a CMake option for release buildsPHASE2_NVTX_RANGES is hard-enabled, adding NVTX instrumentation overhead even in non-profiling builds. The project uses global
RAFT_NVTX ONfor NVTX control; consider tying phase2's local NVTX instrumentation to a CMake option (e.g.,CUOPT_ENABLE_PHASE2_NVTX, defaulting to the RAFT_NVTX setting) so phase-level profiling can be disabled independently when needed in performance-sensitive scenarios.♻️ Suggested approach
-#define PHASE2_NVTX_RANGES +#if defined(RAFT_NVTX) && RAFT_NVTX +#define PHASE2_NVTX_RANGES +#endifOr for finer control, introduce a dedicated CMake option that defaults to RAFT_NVTX.
| struct node_ptr_lower_bound_comp { | ||
| bool operator()(const mip_node_t<i_t, f_t>* a, const mip_node_t<i_t, f_t>* b) const | ||
| { | ||
| return a->lower_bound > b->lower_bound; | ||
| } | ||
| }; | ||
| heap_t<mip_node_t<i_t, f_t>*, node_ptr_lower_bound_comp> heap_; |
There was a problem hiding this comment.
Comparator lacks deterministic tie-breaking for consistent heap ordering.
The node_ptr_lower_bound_comp comparator only compares by lower_bound, but the similar backlog_node_compare_t in bb_worker_state.hpp (lines 46-58) and deterministic_node_ptr_comp in mip_node.hpp (lines 329-340) include BSP identity tuple tie-breaking (origin_worker_id, creation_seq) for determinism. This inconsistency could cause non-deterministic heap ordering when nodes have equal lower bounds.
Suggested fix for deterministic tie-breaking
struct node_ptr_lower_bound_comp {
bool operator()(const mip_node_t<i_t, f_t>* a, const mip_node_t<i_t, f_t>* b) const
{
- return a->lower_bound > b->lower_bound;
+ if (a->lower_bound != b->lower_bound) { return a->lower_bound > b->lower_bound; }
+ // Deterministic tie-breaking by BSP identity tuple
+ if (a->origin_worker_id != b->origin_worker_id) {
+ return a->origin_worker_id > b->origin_worker_id;
+ }
+ return a->creation_seq > b->creation_seq;
}
};🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/branch_and_bound.hpp` around lines 374 - 380, The
comparator node_ptr_lower_bound_comp currently only compares
mip_node_t::lower_bound causing non-deterministic ordering; update operator() in
node_ptr_lower_bound_comp to tie-break when lower_bound are equal by comparing
the BSP identity tuple fields (origin_worker_id then creation_seq) from
mip_node_t<i_t,f_t> so the ordering is deterministic and matches
backlog_node_compare_t / deterministic_node_ptr_comp, and ensure the heap_
instantiation continues to use node_ptr_lower_bound_comp.
| i_t branch_var = variable_selection_from_pseudo_costs(pseudo_cost_sum_down.data(), | ||
| pseudo_cost_sum_up.data(), | ||
| pseudo_cost_num_down.data(), | ||
| pseudo_cost_num_up.data(), | ||
| n, | ||
| fractional, | ||
| solution); | ||
|
|
||
| log.printf( | ||
| "pc branching on %d. Value %e. Score %e\n", branch_var, solution[branch_var], score[select]); | ||
| log.printf("pc branching on %d. Value %e.\n", branch_var, solution[branch_var]); | ||
|
|
There was a problem hiding this comment.
Guard against branch_var == -1 before indexing solution.
variable_selection_from_pseudo_costs returns -1 when fractional is empty. The new log unconditionally indexes solution[branch_var], which becomes UB for -1.
✅ Proposed fix
i_t branch_var = variable_selection_from_pseudo_costs(pseudo_cost_sum_down.data(),
pseudo_cost_sum_up.data(),
pseudo_cost_num_down.data(),
pseudo_cost_num_up.data(),
n,
fractional,
solution);
- log.printf("pc branching on %d. Value %e.\n", branch_var, solution[branch_var]);
+ if (branch_var < 0) { return branch_var; }
+ log.printf("pc branching on %d. Value %e.\n", branch_var, solution[branch_var]);📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| i_t branch_var = variable_selection_from_pseudo_costs(pseudo_cost_sum_down.data(), | |
| pseudo_cost_sum_up.data(), | |
| pseudo_cost_num_down.data(), | |
| pseudo_cost_num_up.data(), | |
| n, | |
| fractional, | |
| solution); | |
| log.printf( | |
| "pc branching on %d. Value %e. Score %e\n", branch_var, solution[branch_var], score[select]); | |
| log.printf("pc branching on %d. Value %e.\n", branch_var, solution[branch_var]); | |
| i_t branch_var = variable_selection_from_pseudo_costs(pseudo_cost_sum_down.data(), | |
| pseudo_cost_sum_up.data(), | |
| pseudo_cost_num_down.data(), | |
| pseudo_cost_num_up.data(), | |
| n, | |
| fractional, | |
| solution); | |
| if (branch_var < 0) { return branch_var; } | |
| log.printf("pc branching on %d. Value %e.\n", branch_var, solution[branch_var]); |
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/pseudo_costs.cpp` around lines 269 - 278, The new log
unconditionally indexes solution[branch_var] which is undefined when
variable_selection_from_pseudo_costs returns -1 (e.g., when fractional is
empty); update the code around branch_var (the call to
variable_selection_from_pseudo_costs and the log.printf) to check if branch_var
== -1 and handle that case (for example, log a message like "pc no branching
variable" or similar) instead of indexing solution, otherwise log the selected
variable and its value as before.
| template <typename i_t, typename f_t> | ||
| pseudo_cost_averages_t<f_t> compute_pseudo_cost_averages( | ||
| const f_t* pc_sum_down, const f_t* pc_sum_up, const i_t* pc_num_down, const i_t* pc_num_up, i_t n) | ||
| { | ||
| i_t num_initialized_down = 0; | ||
| i_t num_initialized_up = 0; | ||
| f_t pseudo_cost_down_avg = 0; | ||
| f_t pseudo_cost_up_avg = 0; | ||
|
|
||
| for (i_t j = 0; j < n; ++j) { | ||
| if (pc_num_down[j] > 0) { | ||
| ++num_initialized_down; | ||
| if (std::isfinite(pc_sum_down[j])) { | ||
| pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j]; | ||
| } | ||
| } | ||
| if (pc_num_up[j] > 0) { | ||
| ++num_initialized_up; | ||
| if (std::isfinite(pc_sum_up[j])) { pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; } | ||
| } | ||
| } | ||
|
|
||
| pseudo_cost_down_avg = | ||
| (num_initialized_down > 0) ? pseudo_cost_down_avg / num_initialized_down : f_t(1.0); | ||
| pseudo_cost_up_avg = | ||
| (num_initialized_up > 0) ? pseudo_cost_up_avg / num_initialized_up : f_t(1.0); | ||
|
|
||
| return {pseudo_cost_down_avg, pseudo_cost_up_avg}; | ||
| } |
There was a problem hiding this comment.
Non‑finite pseudo‑costs skew the computed averages.
num_initialized_* is incremented even when the corresponding sum is non‑finite, which drives averages toward 0 instead of falling back to the intended default when all entries are invalid. This can change branching decisions.
🐛 Proposed fix
for (i_t j = 0; j < n; ++j) {
- if (pc_num_down[j] > 0) {
- ++num_initialized_down;
- if (std::isfinite(pc_sum_down[j])) {
- pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j];
- }
- }
- if (pc_num_up[j] > 0) {
- ++num_initialized_up;
- if (std::isfinite(pc_sum_up[j])) { pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; }
- }
+ if (pc_num_down[j] > 0 && std::isfinite(pc_sum_down[j])) {
+ ++num_initialized_down;
+ pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j];
+ }
+ if (pc_num_up[j] > 0 && std::isfinite(pc_sum_up[j])) {
+ ++num_initialized_up;
+ pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j];
+ }
}Based on learnings, validate algorithm correctness in branching decisions.
📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| template <typename i_t, typename f_t> | |
| pseudo_cost_averages_t<f_t> compute_pseudo_cost_averages( | |
| const f_t* pc_sum_down, const f_t* pc_sum_up, const i_t* pc_num_down, const i_t* pc_num_up, i_t n) | |
| { | |
| i_t num_initialized_down = 0; | |
| i_t num_initialized_up = 0; | |
| f_t pseudo_cost_down_avg = 0; | |
| f_t pseudo_cost_up_avg = 0; | |
| for (i_t j = 0; j < n; ++j) { | |
| if (pc_num_down[j] > 0) { | |
| ++num_initialized_down; | |
| if (std::isfinite(pc_sum_down[j])) { | |
| pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j]; | |
| } | |
| } | |
| if (pc_num_up[j] > 0) { | |
| ++num_initialized_up; | |
| if (std::isfinite(pc_sum_up[j])) { pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; } | |
| } | |
| } | |
| pseudo_cost_down_avg = | |
| (num_initialized_down > 0) ? pseudo_cost_down_avg / num_initialized_down : f_t(1.0); | |
| pseudo_cost_up_avg = | |
| (num_initialized_up > 0) ? pseudo_cost_up_avg / num_initialized_up : f_t(1.0); | |
| return {pseudo_cost_down_avg, pseudo_cost_up_avg}; | |
| } | |
| template <typename i_t, typename f_t> | |
| pseudo_cost_averages_t<f_t> compute_pseudo_cost_averages( | |
| const f_t* pc_sum_down, const f_t* pc_sum_up, const i_t* pc_num_down, const i_t* pc_num_up, i_t n) | |
| { | |
| i_t num_initialized_down = 0; | |
| i_t num_initialized_up = 0; | |
| f_t pseudo_cost_down_avg = 0; | |
| f_t pseudo_cost_up_avg = 0; | |
| for (i_t j = 0; j < n; ++j) { | |
| if (pc_num_down[j] > 0 && std::isfinite(pc_sum_down[j])) { | |
| +num_initialized_down; | |
| pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j]; | |
| } | |
| if (pc_num_up[j] > 0 && std::isfinite(pc_sum_up[j])) { | |
| +num_initialized_up; | |
| pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; | |
| } | |
| } | |
| pseudo_cost_down_avg = | |
| (num_initialized_down > 0) ? pseudo_cost_down_avg / num_initialized_down : f_t(1.0); | |
| pseudo_cost_up_avg = | |
| (num_initialized_up > 0) ? pseudo_cost_up_avg / num_initialized_up : f_t(1.0); | |
| return {pseudo_cost_down_avg, pseudo_cost_up_avg}; | |
| } |
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/pseudo_costs.hpp` around lines 35 - 63, In
compute_pseudo_cost_averages, num_initialized_down/num_initialized_up are
incremented even when pc_sum_down/pc_sum_up is non-finite; change the logic so
you only increment num_initialized_down when pc_num_down[j] > 0 AND
std::isfinite(pc_sum_down[j]) (and likewise only increment num_initialized_up
when pc_num_up[j] > 0 AND std::isfinite(pc_sum_up[j])), leaving the fallback to
f_t(1.0) unchanged when the corresponding count remains zero; update the blocks
around pc_num_down/pc_sum_down and pc_num_up/pc_sum_up so the increment and the
accumulation happen together to avoid skewing averages.
| bool deterministic; // true to use B&B deterministic mode, false to use non-deterministic mode | ||
| bool barrier; // true to use barrier method, false to use dual simplex method |
There was a problem hiding this comment.
The deterministic member is not initialized in the constructor.
The deterministic member (line 143) is declared but not initialized in the constructor initializer list. This leaves it with an indeterminate value (undefined behavior when read). Other boolean members like barrier are initialized in the constructor.
🐛 Proposed fix
Add initialization for deterministic in the constructor initializer list, around line 76 where cudss_deterministic(false) is:
cudss_deterministic(false),
+ deterministic(false),
barrier(false),🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/simplex_solver_settings.hpp` around lines 143 - 144,
Initialize the uninitialized bool member `deterministic` in the
`SimplexSolverSettings` constructor initializer list (the same place where
`cudss_deterministic(false)` is set) to a sensible default (e.g., false) so
`deterministic` does not have an indeterminate value; update the constructor for
the `SimplexSolverSettings` class to include `deterministic(false)` in its
initializer list.
| // Find column singletons | ||
| auto& col_perm_vec = static_cast<std::vector<i_t>&>(col_perm); | ||
| auto& row_perm_vec = static_cast<std::vector<i_t>&>(row_perm); | ||
| row_col_graph_t<i_t> graph{Cdeg.begin(), | ||
| col_perm.begin(), | ||
| A.col_start.cbegin(), | ||
| A.i.cbegin(), | ||
| col_perm_vec.begin(), | ||
| A.col_start.underlying().cbegin(), | ||
| A.i.underlying().cbegin(), | ||
| Rdeg.begin(), | ||
| row_perm.begin(), | ||
| row_perm_vec.begin(), |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
# First, find where VectorI and ins_vector are defined
rg -n "using\s+ins_vector|typedef.*ins_vector|class\s+ins_vector|struct\s+ins_vector" --type cpp --type hpp --type hRepository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
# Search for VectorI definition
rg -n "using\s+VectorI|typedef.*VectorI|class\s+VectorI" --type cpp --type hpp --type hRepository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
# Check the context around lines 203-211 and 237-246 in singletons.cpp
cat -n cpp/src/dual_simplex/singletons.cpp | sed -n '195,255p'Repository: NVIDIA/cuopt
Length of output: 2942
🏁 Script executed:
# Find row_col_graph_t definition to see what template parameters it accepts
rg -n "row_col_graph_t" --type cpp --type hpp --type h -A 5 | head -80Repository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
# Check the function signature containing these lines
rg -n "void.*singletons|row_col_graph_t.*row_perm.*col_perm" cpp/src/dual_simplex/singletons.cpp -B 20 -A 2 | head -150Repository: NVIDIA/cuopt
Length of output: 38
🏁 Script executed:
# Search for VectorI and ins_vector without file type restrictions
rg "using\s+ins_vector|typedef.*ins_vector" -A 2Repository: NVIDIA/cuopt
Length of output: 281
🏁 Script executed:
# Search for VectorI definition
rg "using\s+VectorI|typedef.*VectorI" -A 2Repository: NVIDIA/cuopt
Length of output: 38
🏁 Script executed:
# Find the function containing these lines to see parameter types
rg -B 30 "Find column singletons" cpp/src/dual_simplex/singletons.cpp | head -60Repository: NVIDIA/cuopt
Length of output: 927
🏁 Script executed:
# Search for row_col_graph_t definition
rg "template.*row_col_graph_t|class\s+row_col_graph_t|struct\s+row_col_graph_t" -A 10Repository: NVIDIA/cuopt
Length of output: 1860
🏁 Script executed:
# Check what types row_perm and col_perm are in the function signature
rg "row_perm|col_perm" cpp/src/dual_simplex/singletons.cpp | head -20Repository: NVIDIA/cuopt
Length of output: 1125
🏁 Script executed:
# Find the memop_instrumentation_wrapper_t definition
rg "memop_instrumentation_wrapper_t" -A 10 | head -50Repository: NVIDIA/cuopt
Length of output: 4098
🏁 Script executed:
# Check the function signature more completely
rg -B 5 "VectorI& row_perm" cpp/src/dual_simplex/singletons.cpp | head -30Repository: NVIDIA/cuopt
Length of output: 247
🏁 Script executed:
# Verify the actual template instantiation to see what VectorI resolves to
rg "eliminate_singletons.*ins_vector|eliminate_singletons.*VectorI" -A 2Repository: NVIDIA/cuopt
Length of output: 38
🏁 Script executed:
# Check if memop_instrumentation_wrapper_t has an underlying() method
rg "underlying()" cpp/src/utilities/memory_instrumentation.hpp -B 2 -A 2Repository: NVIDIA/cuopt
Length of output: 8910
🏁 Script executed:
# Get more details on memop_instrumentation_wrapper_t structure
rg "struct memop_instrumentation_wrapper_t" -A 40 cpp/src/utilities/memory_instrumentation.hppRepository: NVIDIA/cuopt
Length of output: 3220
🏁 Script executed:
# Check instantiations of find_singletons to see what VectorI types are used
rg "find_singletons.*std::vector|find_singletons.*ins_vector" -B 2 -A 2Repository: NVIDIA/cuopt
Length of output: 1116
🏁 Script executed:
# Check the actual calls to find_singletons to understand parameter types
rg "find_singletons" --type-add 'cpp:*.cpp' -l | xargs rg -B 5 "find_singletons\(" | head -60Repository: NVIDIA/cuopt
Length of output: 1855
🏁 Script executed:
# Verify row_col_graph_t is NOT templated on iterator type
cat -n cpp/src/dual_simplex/singletons.hpp | sed -n '1,50p'Repository: NVIDIA/cuopt
Length of output: 2110
🏁 Script executed:
# Double-check if there's any std::vector inheritance in memop_instrumentation_wrapper_t
rg "struct memop_instrumentation_wrapper_t.*:.*std::vector"Repository: NVIDIA/cuopt
Length of output: 38
Remove unsafe static_cast for ins_vector type.
The static_cast<std::vector<i_t>&> at lines 204–205 and 237–238 is undefined behavior when VectorI is ins_vector<i_t> (which wraps std::vector<i_t> but does not inherit from it). The cast succeeds at compile-time but fails at runtime due to incompatible memory layout. Use the .underlying() method instead, which safely accesses the wrapped std::vector:
♻️ Suggested fix
- auto& col_perm_vec = static_cast<std::vector<i_t>&>(col_perm);
- auto& row_perm_vec = static_cast<std::vector<i_t>&>(row_perm);
+ auto& col_perm_vec = col_perm.underlying();
+ auto& row_perm_vec = row_perm.underlying();
...
- auto& row_perm_vec2 = static_cast<std::vector<i_t>&>(row_perm);
- auto& col_perm_vec2 = static_cast<std::vector<i_t>&>(col_perm);
+ auto& row_perm_vec2 = row_perm.underlying();
+ auto& col_perm_vec2 = col_perm.underlying();Also applies to: 237–246
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/singletons.cpp` around lines 203 - 211, The code
unsafely uses static_cast<std::vector<i_t>&> on col_perm and row_perm which is
undefined when VectorI is an ins_vector wrapper; replace those casts by calling
the wrapper's .underlying() to get the actual std::vector references (e.g., use
col_perm.underlying() and row_perm.underlying()) before passing into the
row_col_graph_t<i_t> constructor and any other places that currently cast the
vectors (also update the second occurrence around lines 237–246).
There was a problem hiding this comment.
This relies on operator T& in memop_instrumentation_wrapper. for std::vector<> this will be a no-op, for ins_vector this will return the underlying. There is no "runtime failure" since this is a static_cast, not a dynamic_cast. The compilation would fail otherwise.
There was a problem hiding this comment.
@aliceb-nv You're absolutely right, thank you for the clarification! I apologize for the confusion in the original review comment.
The static_cast<std::vector<i_t>&>(col_perm) is indeed well-defined:
- When
VectorIisstd::vector<i_t>, it's an identity cast (no-op) - When
VectorIisins_vector<i_t>, thestatic_castinvokes theoperator T&()conversion operator defined inmemop_instrumentation_wrapper_t, which returns a reference to the underlying vector
This is a valid use of static_cast to invoke a user-defined conversion operator, and there's no undefined behavior. The code is correct as written.
🧠 Learnings used
Learnt from: chris-maes
Repo: NVIDIA/cuopt PR: 500
File: cpp/src/dual_simplex/scaling.cpp:68-76
Timestamp: 2025-12-04T04:11:12.640Z
Learning: In the cuOPT dual simplex solver, CSR/CSC matrices (including the quadratic objective matrix Q) are required to have valid dimensions and indices by construction. Runtime bounds checking in performance-critical paths like matrix scaling is avoided to prevent slowdowns. Validation is performed via debug-only check_matrix() calls wrapped in `#ifdef` CHECK_MATRIX.
Learnt from: chris-maes
Repo: NVIDIA/cuopt PR: 746
File: cpp/src/dual_simplex/barrier.cu:549-552
Timestamp: 2026-01-14T00:38:38.038Z
Learning: In cpp/src/dual_simplex/barrier.cu's form_adat() method within the barrier solver, the raft::copy from d_original_A_values to device_AD.x is necessary and must not be removed. This copy restores the original matrix A values before they are scaled by the diagonal matrix D (via d_inv_diag_prime). Since D changes between iterations and the scaling mutates device_AD.x, copying the original values each iteration is required for correctness.
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Identify assertions with overly strict numerical tolerances that fail on legitimate degenerate/edge cases (near-zero pivots, singular matrices, empty problems)
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Validate algorithm correctness in optimization logic: simplex pivots, branch-and-bound decisions, routing heuristics, and constraint/objective handling must produce correct results
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Validate correct initialization of variable bounds, constraint coefficients, and algorithm state before solving; ensure reset when transitioning between algorithm phases (presolve, simplex, diving, crossover)
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Ensure variables and constraints are accessed from the correct problem context (original vs presolve vs folded vs postsolve); verify index mapping consistency across problem transformations
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh} : Avoid reinventing functionality already available in Thrust, CCCL, or RMM libraries; prefer standard library utilities over custom implementations
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Refactor code duplication in solver components (3+ occurrences) into shared utilities; for GPU kernels, use templated device functions to avoid duplication
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Assess algorithmic complexity for large-scale problems (millions of variables/constraints); ensure O(n log n) or better complexity, not O(n²) or worse
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*test*.{cpp,cu,py} : Test with free variables, singleton problems, and extreme problem dimensions near resource limits to validate edge case handling
| typename model_t::Entry data[model_t::NUM_FEATURES]; | ||
| for (int i = 0; i < model_t::NUM_FEATURES; ++i) { | ||
| if (features.find(std::string(model_t::feature_names[i])) == features.end()) { | ||
| data[i].missing = -1; | ||
| CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]); | ||
| } else { | ||
| data[i].fvalue = features.at(std::string(model_t::feature_names[i])); | ||
| } | ||
| } |
There was a problem hiding this comment.
Union member access after setting different member is undefined behavior.
When a feature is missing, data[i].missing = -1 is assigned. When present, data[i].fvalue is assigned. Later at line 57, data[i].missing != -1 is checked regardless of which member was last written. In C++, reading a union member that wasn't the last one written is undefined behavior (except for common initial sequences or type-punning through char*).
Additionally, the map is looked up twice unnecessarily (once with find, once with at).
Proposed fix
typename model_t::Entry data[model_t::NUM_FEATURES];
+ bool is_missing[model_t::NUM_FEATURES];
for (int i = 0; i < model_t::NUM_FEATURES; ++i) {
- if (features.find(std::string(model_t::feature_names[i])) == features.end()) {
+ auto it = features.find(std::string(model_t::feature_names[i]));
+ if (it == features.end()) {
data[i].missing = -1;
+ is_missing[i] = true;
CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]);
} else {
- data[i].fvalue = features.at(std::string(model_t::feature_names[i]));
+ data[i].fvalue = it->second;
+ is_missing[i] = false;
}
}
std::vector<float> cache_vec;
cache_vec.reserve(model_t::NUM_FEATURES);
for (int i = 0; i < model_t::NUM_FEATURES; ++i) {
- cache_vec.push_back(data[i].missing != -1 ? data[i].fvalue
+ cache_vec.push_back(!is_missing[i] ? static_cast<float>(data[i].fvalue)
: std::numeric_limits<float>::quiet_NaN());
}🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_predictor.cpp` around lines 44 - 52, The code
writes different union members of model_t::Entry (data[]) depending on presence
of a feature then later reads the other member, causing undefined behavior;
change model_t::Entry to include an explicit tag/flag (e.g., bool is_missing or
an enum) or replace the union with a struct containing std::optional<float> so
you always have a well-defined way to test presence, set that flag/optional when
populating data[i] (use the iterator returned by features.find to avoid calling
features.at twice) and update the later check (the place currently checking
data[i].missing != -1) to use the new flag/optional instead of reading a
possibly-uninitialized union member; ensure all code paths that construct or
read model_t::Entry use the new presence indicator.
| uint32_t key = cuopt::linear_programming::detail::compute_hash(cache_vec); | ||
|
|
||
| auto cached_it = prediction_cache.find(key); | ||
| if (cached_it != prediction_cache.end()) { return cached_it->second; } |
There was a problem hiding this comment.
Hash collision can return incorrect cached predictions.
Using a 32-bit hash as the direct cache key without storing or verifying the original feature values means hash collisions will silently return incorrect cached results. This could cause subtle bugs and non-determinism.
Consider either:
- Storing the full feature vector alongside the cached result for verification
- Using a collision-resistant key (e.g., the feature vector itself via
std::map)
Proposed fix using full key verification
- auto cached_it = prediction_cache.find(key);
- if (cached_it != prediction_cache.end()) { return cached_it->second; }
+ auto cached_it = prediction_cache.find(key);
+ if (cached_it != prediction_cache.end() && cached_it->second.first == cache_vec) {
+ return cached_it->second.second;
+ }
// ... after prediction ...
- prediction_cache[key] = scaled_result;
+ prediction_cache[key] = {cache_vec, scaled_result};This requires changing the cache type to store std::pair<std::vector<float>, float>.
🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_predictor.cpp` around lines 60 - 63, The code
uses a 32-bit key from
cuopt::linear_programming::detail::compute_hash(cache_vec) to index
prediction_cache, which allows hash collisions to return wrong predictions;
change the cache to store and verify the full feature vector with each cached
value (e.g., map key -> pair<vector<float>, float> or store vector in the mapped
value) and on lookup (when computing key and finding cached_it) compare
cached_it->second's stored vector to cache_vec and only return the cached
prediction on exact match, otherwise treat as a miss and compute/store a new
entry; update any insertion logic to save the full cache_vec alongside the
prediction and adjust prediction_cache's type accordingly (alternatively change
the container to use the feature vector itself as the key).
There was a problem hiding this comment.
Extremely unlikely to occur, and if it does the consequences are minor (incorrect work unit count, delayed sync). Will be reworked later in either case with the GPU determinism
There was a problem hiding this comment.
@aliceb-nv Understood, thanks for the context. Given the low likelihood and minor impact, plus the upcoming GPU determinism rework, this approach sounds reasonable. I'll move on from this comment.
🧠 Learnings used
Learnt from: aliceb-nv
Repo: NVIDIA/cuopt PR: 798
File: cpp/src/utilities/models/fj_predictor/quantize.cpp:1149-1161
Timestamp: 2026-01-26T14:17:20.839Z
Learning: The files in cpp/src/utilities/models/*/quantize.cpp (fj_predictor, cpufj_predictor, pdlp_predictor, dualsimplex_predictor, bounds_strengthening_predictor) are autogenerated model artifacts and should not have manual code changes suggested; any improvements should target the model generation tooling or calling code validation layer.
Learnt from: aliceb-nv
Repo: NVIDIA/cuopt PR: 798
File: cpp/src/utilities/models/pdlp_predictor/quantize.cpp:975-987
Timestamp: 2026-01-26T14:17:22.967Z
Learning: Files in cpp/src/utilities/models/*/quantize.cpp (and related model predictor files like header.h, main.cpp) are autogenerated and should not receive suggestions for manual code changes. These are ML model artifacts (e.g., fj_predictor, cpufj_predictor, dualsimplex_predictor, pdlp_predictor, bounds_strengthening_predictor).
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Verify correct problem size checks before expensive GPU/CPU operations; prevent resource exhaustion on oversized problems
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Ensure race conditions are absent in multi-GPU code and multi-threaded server implementations; verify proper synchronization of shared state
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Refactor code duplication in solver components (3+ occurrences) into shared utilities; for GPU kernels, use templated device functions to avoid duplication
| // Temporary scaling classes until I figure out better ways to do this | ||
| // to account for performance differences between the regression learning machine and the user | ||
| // machine. (e.g. integrate memory latency/bandwidth, cache topology, user-provided tuning...) | ||
| struct cpu_work_unit_scaler_t { | ||
| cpu_work_unit_scaler_t() | ||
| { | ||
| constexpr double baseline_max_clock = 3800.0; | ||
| double max_clock = get_cpu_max_clock_mhz(); | ||
| scaling_factor_ = baseline_max_clock / max_clock; | ||
| } | ||
|
|
||
| double scale_work_units(double work_units) const { return work_units * scaling_factor_; } | ||
|
|
||
| private: | ||
| double scaling_factor_; | ||
| }; |
There was a problem hiding this comment.
Guard against zero/invalid CPU max clock.
If get_cpu_max_clock_mhz() returns 0/negative, Line 37 divides by zero and propagates inf/NaN scaling.
🐛 Suggested guard
- double max_clock = get_cpu_max_clock_mhz();
- scaling_factor_ = baseline_max_clock / max_clock;
+ const double max_clock = get_cpu_max_clock_mhz();
+ scaling_factor_ =
+ (max_clock > 0.0) ? (baseline_max_clock / max_clock) : 1.0;🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_predictor.hpp` around lines 29 - 44, The
constructor for cpu_work_unit_scaler_t divides baseline_max_clock by
get_cpu_max_clock_mhz() without validating the result, which can produce inf/NaN
if max_clock <= 0 or non-finite; update the constructor to validate max_clock
(call get_cpu_max_clock_mhz()), check for <= 0 or non-finite (use std::isfinite
or equivalent) and if invalid set scaling_factor_ to a safe default (e.g., 1.0)
or compute using a clamped minimum value, otherwise compute scaling_factor_ =
baseline_max_clock / max_clock; ensure scaling_factor_ is always initialized to
a finite value.
| double work_unit_scheduler_t::current_sync_target() const | ||
| { | ||
| if (sync_interval_ <= 0) return std::numeric_limits<double>::infinity(); | ||
| return (barrier_generation_ + 1) * sync_interval_; | ||
| } |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
find . -name "work_unit_scheduler.*" -type fRepository: NVIDIA/cuopt
Length of output: 143
🏁 Script executed:
fd -e h -e hpp -e cpp "work_unit_scheduler" --exec cat {} \;Repository: NVIDIA/cuopt
Length of output: 9679
🏁 Script executed:
rg "barrier_generation_" -A 3 -B 3 --type-list | head -20Repository: NVIDIA/cuopt
Length of output: 445
🏁 Script executed:
rg "class work_unit_scheduler" -A 30Repository: NVIDIA/cuopt
Length of output: 2416
🏁 Script executed:
rg "barrier_generation_" cpp/src/utilities/ -A 2 -B 2Repository: NVIDIA/cuopt
Length of output: 2612
Guard barrier_generation_ reads against concurrent modification by writer threads.
current_sync_target() reads barrier_generation_ without synchronization while wait_at_sync_point() modifies it under mutex protection, creating a data race. This undefined behavior must be prevented. Additionally, on_work_recorded() reads barrier_generation_ directly in logging without synchronization.
Acquire the mutex before reading barrier_generation_ in current_sync_target(), or alternatively declare the member as std::atomic<size_t> if atomic semantics are intended.
🔧 Suggested fix
double work_unit_scheduler_t::current_sync_target() const
{
if (sync_interval_ <= 0) return std::numeric_limits<double>::infinity();
+ std::lock_guard<std::mutex> lock(mutex_);
return (barrier_generation_ + 1) * sync_interval_;
}📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| double work_unit_scheduler_t::current_sync_target() const | |
| { | |
| if (sync_interval_ <= 0) return std::numeric_limits<double>::infinity(); | |
| return (barrier_generation_ + 1) * sync_interval_; | |
| } | |
| double work_unit_scheduler_t::current_sync_target() const | |
| { | |
| if (sync_interval_ <= 0) return std::numeric_limits<double>::infinity(); | |
| std::lock_guard<std::mutex> lock(mutex_); | |
| return (barrier_generation_ + 1) * sync_interval_; | |
| } |
🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_scheduler.cpp` around lines 103 - 107,
current_sync_target() and on_work_recorded() read barrier_generation_ without
synchronization while wait_at_sync_point() mutates it under a mutex, causing a
data race; fix by either acquiring the same mutex (the mutex used in
wait_at_sync_point()) around reads in current_sync_target() and where
on_work_recorded() logs barrier_generation_, or convert barrier_generation_ to
std::atomic<size_t> and use atomic reads; pick one approach and update all read
sites (current_sync_target(), on_work_recorded()) to use the chosen
synchronization so reads are safe against concurrent writers.
📝 WalkthroughWalkthroughAdds deterministic mode support for MIP solving with work-limit enforcement and fixed seeding. Introduces memory instrumentation framework for tracking memory access, bulk synchronous parallel (BSP) branch-and-bound execution, work-unit prediction models, and extensive instrumentation infrastructure across the dual simplex solver. Changes
Estimated code review effort🎯 4 (Complex) | ⏱️ ~75 minutes 🚥 Pre-merge checks | ✅ 1 | ❌ 2❌ Failed checks (1 warning, 1 inconclusive)
✅ Passed checks (1 passed)
✏️ Tip: You can configure your own custom pre-merge checks in the settings. ✨ Finishing touches
Thanks for using CodeRabbit! It's free for OSS, and your support helps us grow. If you like it, consider giving us a shout-out. Comment |
There was a problem hiding this comment.
Actionable comments posted: 13
Note
Due to the large number of review comments, Critical, Major severity comments were prioritized as inline comments.
Caution
Some comments are outside the diff and can’t be posted inline due to platform limitations.
⚠️ Outside diff range comments (3)
cpp/src/dual_simplex/solve.hpp (1)
34-48: MissingWORK_LIMITcase inlp_status_to_string.The switch statement handles the new
UNSETstatus but does not handleWORK_LIMIT. This will causelp_status_to_string(lp_status_t::WORK_LIMIT)to return"UNKNOWN", which will be confusing in logs and diagnostics.Suggested fix
case lp_status_t::CONCURRENT_LIMIT: return "CONCURRENT_LIMIT"; case lp_status_t::UNSET: return "UNSET"; + case lp_status_t::WORK_LIMIT: return "WORK_LIMIT"; } return "UNKNOWN";cpp/src/dual_simplex/simplex_solver_settings.hpp (1)
42-78: Initializedeterministicbool member in constructor initializer list.The
deterministicmember variable declared at line 143 is not initialized in the constructor, leaving it with an indeterminate value. This can randomly enable or disable B&B deterministic mode.🛠️ Proposed fix
barrier_presolve(false), cudss_deterministic(false), + deterministic(false), barrier(false),Also applies to: 143-144
cpp/src/dual_simplex/branch_and_bound.cpp (1)
524-556: Avoid multipleheuristic_preemption_callbackinvocations.The callback is invoked unconditionally near the top and then again in several termination branches. If the callback isn’t idempotent, this can double-stop threads or race with cleanup.
🛠️ Suggested fix (call once after final status is known)
- if (settings_.heuristic_preemption_callback != nullptr) { - settings_.heuristic_preemption_callback(); - } + bool preempted = false; + auto preempt_once = [&]() { + if (!preempted && settings_.heuristic_preemption_callback != nullptr) { + settings_.heuristic_preemption_callback(); + preempted = true; + } + }; ... - settings_.heuristic_preemption_callback(); + preempt_once(); ... - settings_.heuristic_preemption_callback(); + preempt_once(); ... + preempt_once(); return solver_status_;
🤖 Fix all issues with AI agents
In `@cpp/src/dual_simplex/branch_and_bound.cpp`:
- Around line 2713-2745: process_history_and_sync() currently ignores
repair_queue_ (TODO commented) so infeasible solutions enqueued by
set_new_solution_deterministic() are never processed; fix by draining
repair_queue_ into the local to_repair vector at the start of the repair block:
lock mutex_repair_, move or swap repair_queue_ into to_repair, clear
repair_queue_, unlock mutex_repair_, then proceed with the existing sort/repair
loop and push repaired entries into heuristic_solution_queue_ (using
mutex_heuristic_queue_ and bsp_current_horizon_); ensure the move preserves
deterministic ordering (e.g., preserve insertion order before sorting) so
deterministic runs actually repair queued solutions.
- Around line 3041-3066: compute_bsp_lower_bound() only reads per-worker nodes
and can return inf because it omits the global lower_bound_ceiling_; include
lower_bound_ceiling_ in the computed minimum so numerical ceilings are enforced.
Update the function (branch_and_bound_t<i_t, f_t>::compute_bsp_lower_bound) to
initialize lower_bound to std::min(inf, lower_bound_ceiling_) or explicitly take
lower_bound = lower_bound_ceiling_ and then min with
worker.current_node->lower_bound, worker.plunge_stack nodes, and
worker.backlog.data() nodes before returning.
- Around line 1869-1883: BSP path currently unconditionally populates
diving_types with all four strategies; instead, filter by the per-heuristic
enable flags in settings before calling bsp_diving_workers_->initialize. In the
block that constructs diving_types (used by bsp_diving_workers_), check each
corresponding settings flag (e.g., the settings_ flags that control pseudocost,
line-search, guided and coefficient diving) and push only enabled
bnb_worker_type_t values; if the resulting vector is empty, skip
creating/initializing bsp_diving_workers_ so BSP remains consistent with the
non-BSP behavior. Ensure you still call calculate_variable_locks(original_lp_,
var_up_locks_, var_down_locks_) only when workers are actually initialized.
In `@cpp/src/dual_simplex/phase2.cpp`:
- Around line 2486-2492: The code registers A_transpose into the instrumentation
manifold twice via calls to
manifold.add("A_transpose.col_start"/"A_transpose.i"/"A_transpose.x") both
before and after lp.A.transpose(A_transpose); remove the duplicate registration
so each A_transpose.* key is added only once (choose whether to register before
or after transpose and delete the other set), and apply the same
single-registration fix to the other duplicate group around the second transpose
(the block at the later occurrence). Ensure only one set of manifold.add calls
references A_transpose.col_start, A_transpose.i, and A_transpose.x.
In `@cpp/src/mip/diversity/diversity_manager.cu`:
- Around line 309-345: The current deterministic and "heuristics disabled"
branches spin on check_b_b_preemption() with no timeout and no null-B&B
fallback, which can block forever; update both loops to enforce a maximum wait
(e.g., derive from a setting or use a reasonable default) and break if the
timeout elapses, call ls.stop_cpufj_deterministic() if CPUFJ was started, and if
context.branch_and_bound_ptr is null skip waiting entirely and return
population.best_feasible() immediately; make these changes around the existing
symbols population.initialize_population(), population.allocate_solutions(),
ls.start_cpufj_deterministic(), ls.stop_cpufj_deterministic(), and
population.best_feasible() so the code never hangs waiting for B&B.
In `@cpp/src/utilities/hashing.hpp`:
- Around line 8-20: The header is missing the <cstring> include required for
std::memcpy used in compute_hash (and the device/host code paths that call it);
add `#include` <cstring> at the top of the file so std::memcpy is declared and
builds do not rely on transitive includes for functions used in compute_hash and
any other functions in this header that call memcpy.
In `@cpp/src/utilities/models/cpufj_predictor/quantize.cpp`:
- Around line 262-274: The function quantize accesses th_begin[fid],
th_len[fid], and threshold without validating fid; add a bounds check at the
start of cpufj_predictor::quantize to ensure fid is less than the size of
th_begin/th_len (e.g., num_features or th_begin.size()), and if out of range
return the same sentinel used elsewhere (e.g., -10) or handle as appropriate;
update any early-return condition that currently compares offset to a hardcoded
value (offset == 231) to instead rely on the validated offset derived from
th_begin[fid].
In `@cpp/src/utilities/producer_sync.hpp`:
- Around line 19-23: The header is missing <algorithm>, which is required for
std::find used in deregister_producer(); add `#include` <algorithm> to the
includes at the top of producer_sync.hpp so deregister_producer() (and any other
uses of std::find) will compile correctly.
In `@cpp/src/utilities/work_unit_predictor.cpp`:
- Around line 44-52: The loop currently leaves model_t::Entry::missing
uninitialized for present features causing nondeterministic behavior; in the
else branch where you set data[i].fvalue (and when features.at(...) succeeds),
also explicitly set data[i].missing = 0 (and initialize any other Entry fields
that must have deterministic defaults) so every data[i] has a defined missing
flag; update the block that references model_t::feature_names, features.at(...),
and CUOPT_LOG_WARN to set missing=0 in the present-feature path.
- Around line 62-74: prediction_cache is accessed concurrently by the static
predictor instance causing a data race; add a mutex (e.g., std::mutex
prediction_cache_mutex as a member of the predictor class) and protect all
accesses to prediction_cache with a lock (use std::lock_guard<std::mutex>
guard(prediction_cache_mutex)) around the find/return path and around the code
that assigns prediction_cache[key] = scaled_result; keep model_t::predict and
scaler_.scale_work_units calls outside the locked region if possible (or
document if they must be inside), ensuring both lookup and insertion are
synchronized to eliminate the race.
In `@cpp/src/utilities/work_unit_predictor.hpp`:
- Around line 33-38: The cpu_work_unit_scaler_t constructor uses
get_cpu_max_clock_mhz() to compute scaling_factor_, but if max_clock is 0 or
negative the result becomes INF; update the constructor (cpu_work_unit_scaler_t)
to guard against invalid readings by checking if max_clock <= 0 and in that case
set a safe fallback (e.g., set scaling_factor_ = 1.0 or use baseline_max_clock
as the divisor) and optionally log the anomaly; ensure scaling_factor_ is always
finite before storing it.
- Around line 58-60: The prediction_cache is not thread-safe; add a mutable
std::mutex member (e.g., prediction_cache_mutex) and `#include` <mutex>, then
protect all accesses to prediction_cache inside predict_scalar() with a lock
(use std::lock_guard<std::mutex> or similar) around the find/read and
insert/update paths so concurrent calls from OpenMP threads cannot race while
still leaving scaler_ untouched.
In `@cpp/tests/mip/determinism_test.cu`:
- Around line 68-71: The parameterized test (DeterministicBBInstanceTest) calls
cuopt::seed_generator::set_seed(seed) but fails to set settings.seed, so
downstream code using settings.seed sees -1; before invoking solve_mip() in that
test, assign settings.seed = seed to match non-parameterized tests and ensure
all RNG initializations (e.g., those reading settings.seed and the solver's seed
setup) receive the intended seed value.
🟡 Minor comments (18)
cpp/src/mip/feasibility_jump/fj_cpu.cu-397-401 (1)
397-401: Division by zero if time_window_ms is zero.
mem_bandwidth_gb_per_seccalculation can divide by zero iftime_window_msis zero (e.g., on first call or very fast execution).🐛 Proposed fix
- double mem_bandwidth_gb_per_sec = (mem_total_mb / 1000.0) / (time_window_ms / 1000.0); + double mem_bandwidth_gb_per_sec = time_window_ms > 0.0 + ? (mem_total_mb / 1000.0) / (time_window_ms / 1000.0) + : 0.0;cpp/src/mip/feasibility_jump/fj_cpu.cu-296-341 (1)
296-341: Potential division by zero in precompute_problem_features.If
n_varsorn_cstrsis zero, lines 316, 335, and 340 will divide by zero. While empty problems are rare, defensive checks would improve robustness.🐛 Proposed fix
i_t total_nnz = fj_cpu.h_reverse_offsets.back(); i_t n_vars = fj_cpu.h_reverse_offsets.size() - 1; i_t n_cstrs = fj_cpu.h_offsets.size() - 1; + if (n_vars == 0 || n_cstrs == 0) { + // Empty problem - set defaults and return early + fj_cpu.avg_var_degree = 0.0; + fj_cpu.avg_cstr_degree = 0.0; + fj_cpu.var_degree_cv = 0.0; + fj_cpu.cstr_degree_cv = 0.0; + fj_cpu.problem_density = 0.0; + return; + } + fj_cpu.avg_var_degree = (double)total_nnz / n_vars;cpp/src/utilities/version_info.cpp-166-201 (1)
166-201: Minor:std::stodcan throw on malformed input.If
/proc/cpuinfocontains a malformed "cpu MHz" line,std::stodat line 191 will throwstd::invalid_argumentorstd::out_of_range, causing the lambda (and thus the first call toget_cpu_max_clock_mhz()) to propagate an exception.Consider wrapping the parse in a try-catch or using a safer parsing approach:
Suggested fix
if (line.find("cpu MHz") != std::string::npos) { std::size_t colon = line.find(':'); if (colon != std::string::npos) { - double mhz = std::stod(line.substr(colon + 1)); - if (mhz > max_mhz) { max_mhz = mhz; } + try { + double mhz = std::stod(line.substr(colon + 1)); + if (mhz > max_mhz) { max_mhz = mhz; } + } catch (...) { + // Skip malformed lines + } } }cpp/src/utilities/timing_utils.hpp-18-56 (1)
18-56: Use PRIu64 for portableuint64_tprinting.
%luis incorrect on LLP64 platforms like Windows, whereunsigned longis 32-bit. UsingPRIu64from<cinttypes>ensures the correct platform-specific format specifier and prevents truncation.🛠️ Proposed fix
`#include` <algorithm> `#include` <array> `#include` <cstdint> +#include <cinttypes> `#include` <cstdio> @@ - printf(`#name` ": avg=%lu cycles, median=%lu cycles (n=%zu)\n", \ - avg_##name, median_##name, timing_buffer_size_##name); \ + printf(`#name` ": avg=%" PRIu64 " cycles, median=%" PRIu64 " cycles (n=%zu)\n", \ + avg_##name, median_##name, timing_buffer_size_##name); \cpp/src/dual_simplex/pseudo_costs.hpp-44-60 (1)
44-60: Don’t count non‑finite pseudo‑costs in the average.If a sum is non‑finite, you currently still increment the initialized count, which biases the average downward. Consider only counting when the sum is finite.
🐛 Suggested adjustment
- if (pc_num_down[j] > 0) { - ++num_initialized_down; - if (std::isfinite(pc_sum_down[j])) { - pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j]; - } - } + if (pc_num_down[j] > 0 && std::isfinite(pc_sum_down[j])) { + ++num_initialized_down; + pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j]; + } if (pc_num_up[j] > 0) { - ++num_initialized_up; - if (std::isfinite(pc_sum_up[j])) { pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; } + if (std::isfinite(pc_sum_up[j])) { + ++num_initialized_up; + pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; + } }cpp/src/dual_simplex/mip_node.hpp-311-321 (1)
311-321: Tie‑breaker comparison contradicts the comment.The code says “prefer deeper nodes,” but the comparator currently makes deeper nodes lower priority. Either flip the comparison or update the comment to match the intended behavior.
🛠️ Suggested fix (if deeper nodes should win ties)
- return a.depth > b.depth; + return a.depth < b.depth; ... - return a->depth > b->depth; + return a->depth < b->depth;cpp/src/mip/problem/problem_helpers.cuh-359-360 (1)
359-360: Keep CSR validation for empty matrices.
The early return skipscheck_csr_representation, so malformed offsets/indices won’t be caught in debug even whennnz == 0. Consider validating before returning.Proposed fix
- // skip if the matrix is empty - if (values.size() == 0) { return; } + // skip if the matrix is empty (but still validate CSR invariants) + if (values.size() == 0) { + check_csr_representation(values, offsets, indices, handle_ptr, cols, rows); + return; + }cpp/src/dual_simplex/node_queue.hpp-62-64 (1)
62-64: Theheap_t::data()method is not explicitly protected and returns an unordered buffer; however, review the architectural synchronization model.The underlying heap buffer is indeed unordered (a partial heap structure, not sorted), and the method lacks internal locking. However, all five call sites in the codebase occur within
bsp_sync_callback()or functions invoked from it—which execute after workers pause at the BSP synchronization barrier. At these points, no concurrent access is possible. If future code accessesheap_t::data()outside the BSP sync phase, explicit synchronization will be needed; ensure any new uses are either protected by a lock or documented as BSP-sync-only.cpp/src/dual_simplex/solve.cpp-194-205 (1)
194-205: NVTX scope label mismatch: Phase 1 code labeled as "phase2".The NVTX range on line 195 is labeled
"DualSimplex::phase2"but it wraps the Phase 1 solve call (dual_phase2(1, ...)where the first argument1indicates phase 1). This will cause confusion when analyzing profiler traces.Suggested fix
- raft::common::nvtx::range scope_phase1("DualSimplex::phase2"); + raft::common::nvtx::range scope_phase1("DualSimplex::phase1");cpp/include/cuopt/linear_programming/mip/solver_settings.hpp-81-83 (1)
81-83: Add Doxygen documentation forwork_limitto complete public API docs.
determinism_mode(lines 95–103) andseed(lines 106–110) have comprehensive Doxygen documentation, butwork_limit(line 82) lacks any comment. Add a brief doc comment to maintain consistency in the public API header.📝 Suggested doc stub
f_t time_limit = std::numeric_limits<f_t>::infinity(); + /** + * `@brief` Work limit for deterministic runs (work units). Use infinity to disable. + */ f_t work_limit = std::numeric_limits<f_t>::infinity();cpp/src/utilities/models/dualsimplex_predictor/quantize.cpp-1860-1873 (1)
1860-1873: Handleth_len == 0to avoid quantizing unused features.
Some entries inth_lenare zero whileth_beginis not sentinel; the current path can return bins from unrelated thresholds.🛠️ Suggested fix
int dualsimplex_predictor::quantize(float val, unsigned fid) { const size_t offset = th_begin[fid]; const float* array = &threshold[offset]; int len = th_len[fid]; int low = 0; int high = len; int mid; float mval; // It is possible th_begin[i] == [total_num_threshold]. This means that // all features i, (i+1), ... are not used for any of the splits in the model. // So in this case, just return something - if (offset == 1801 || val < array[0]) { return -10; } + if (len == 0 || offset == 1801 || val < array[0]) { return -10; }cpp/tests/linear_programming/c_api_tests/c_api_test.c-1205-1220 (1)
1205-1220: Guard againstnum_runs <= 0to avoid a false pass.
Whennum_runsis zero, the loop never executes and the test reports success without validating determinism.🛠️ Suggested fix
printf("Testing deterministic B&B: %s with %d threads, %d runs\n", filename, num_threads, num_runs); + if (num_runs <= 0) { + printf("Error: num_runs must be > 0\n"); + status = CUOPT_INVALID_ARGUMENT; + goto DONE; + } + status = cuOptReadProblem(filename, &problem);cpp/src/mip/solve.cu-144-146 (1)
144-146: Unsafe use ofstd::getenv- potential null dereference and use ofatoi.If
CUOPT_MIP_HIDE_SOLUTIONis not set,std::getenvreturnsnullptr. While the ternary guards against passingnullptrtoatoi, callingstd::getenvtwice is redundant. Also,atoidoesn't provide error handling for invalid input.🛡️ Proposed fix
- int hidesol = - std::getenv("CUOPT_MIP_HIDE_SOLUTION") ? atoi(std::getenv("CUOPT_MIP_HIDE_SOLUTION")) : 0; + const char* hidesol_env = std::getenv("CUOPT_MIP_HIDE_SOLUTION"); + int hidesol = hidesol_env ? std::atoi(hidesol_env) : 0; if (!hidesol) { detail::print_solution(scaled_problem.handle_ptr, sol.get_solution()); }cpp/tests/mip/determinism_test.cu-101-104 (1)
101-104: Remove extraneous semicolon.Line 102 has a stray semicolon after the seed initialization.
🔧 Proposed fix
auto seed = std::random_device{}() & 0x7fffffff; - ; std::cout << "Tested with seed " << seed << "\n";cpp/tests/mip/determinism_test.cu-159-161 (1)
159-161: Remove extraneous semicolon.Same issue as above - stray semicolon on line 160.
🔧 Proposed fix
auto seed = std::random_device{}() & 0x7fffffff; - ; std::cout << "Tested with seed " << seed << "\n";cpp/src/utilities/work_limit_timer.hpp-133-140 (1)
133-140: Potential null pointer dereference inelapsed_time().In deterministic mode,
elapsed_time()accesseswork_context->global_work_units_elapsedwithout checking ifwork_contextis null, unlike other methods that include null checks.Proposed fix
double elapsed_time() const noexcept { if (deterministic) { + if (!work_context) { return 0.0; } return work_context->global_work_units_elapsed - work_units_at_start; } else { return timer.elapsed_time(); } }cpp/src/mip/problem/problem.cu-881-928 (1)
881-928: Potential division by zero in unbalancedness calculation.When
meanis zero (e.g., for a problem where all rows have zero non-zeros, though unlikely), line 926 will cause a division by zero.Proposed fix
nnz_stddev = std::sqrt(variance); - unbalancedness = nnz_stddev / mean; + unbalancedness = (mean > 0.0) ? nnz_stddev / mean : 0.0; }cpp/src/utilities/memory_instrumentation.hpp-54-76 (1)
54-76: Thread safety: counters are not synchronized.The
byte_loadsandbyte_storescounters are modified byrecord_load(),record_store(), andrecord_rmw()without synchronization. If anins_vectorinstance is accessed from multiple threads concurrently, this causes a data race.The BSP design appears to use per-worker containers (each
bb_worker_state_towns its ownleaf_problem), which would avoid this issue. However, consider adding a comment documenting that instrumented containers must not be shared across threads, or use atomics if shared access is intended.
🧹 Nitpick comments (32)
cpp/src/dual_simplex/vector_math.hpp (1)
86-95: Consider consistent size handling across permutation utilities.This function resizes
pinvif the size doesn't match, whilepermute_vectorandinverse_permute_vectorassert on size mismatch. The different behaviors may be intentional (this function creates new data vs. transforming existing), but the inconsistency could confuse maintainers.Consider either:
- Adding a comment explaining the intentional difference, or
- Aligning behavior (e.g., always assert, or always resize for output vectors)
cpp/src/mip/feasibility_jump/fj_cpu.cuh (1)
155-158: Consider documenting the work_unit_bias semantics.The
work_unit_biasfield defaults to0.8with a comment about keeping CPUFJ ahead of B&B. A brief note explaining why values< 1.0make CPUFJ appear faster would improve maintainability.cpp/src/mip/feasibility_jump/fj_cpu.cu (2)
100-105: Unused variablenew_valin get_mtm_for_constraint.
new_valis declared and assigned on line 102 but never used. This appears to be dead code.♻️ Remove unused variable
// skip constraints that are violated/satisfied based on the MTM move type bool violated = slack < -cstr_tolerance; if (move_type == MTMMoveType::FJ_MTM_VIOLATED ? !violated : violated) continue; - f_t new_val = old_val; - delta_ij = slack / (cstr_coeff * sign); break;
893-906: Variable shadowing:c_lb,c_ub, anddeltashadow outer declarations.Lines 895-897 declare
c_lb,c_ub, anddeltawhich shadow the variables declared at lines 854-856 and 885. This can lead to confusion and subtle bugs.♻️ Rename inner variables to avoid shadowing
} else { auto cstr_coeff = fj_cpu.h_coefficients[i]; - f_t c_lb = fj_cpu.h_cstr_lb[cstr_idx]; - f_t c_ub = fj_cpu.h_cstr_ub[cstr_idx]; - auto [delta, sign, slack, cstr_tolerance] = + f_t inner_c_lb = fj_cpu.h_cstr_lb[cstr_idx]; + f_t inner_c_ub = fj_cpu.h_cstr_ub[cstr_idx]; + auto [mtm_delta, sign, slack, mtm_cstr_tolerance] = get_mtm_for_constraint<i_t, f_t, move_type>(fj_cpu.view, var_idx, cstr_idx, cstr_coeff, - c_lb, - c_ub, + inner_c_lb, + inner_c_ub, fj_cpu.h_assignment, fj_cpu.h_lhs); if (is_integer_var<i_t, f_t>(fj_cpu, var_idx)) { new_val = cstr_coeff * sign > 0 - ? floor(val + delta + fj_cpu.view.pb.tolerances.integrality_tolerance) - : ceil(val + delta - fj_cpu.view.pb.tolerances.integrality_tolerance); + ? floor(val + mtm_delta + fj_cpu.view.pb.tolerances.integrality_tolerance) + : ceil(val + mtm_delta - fj_cpu.view.pb.tolerances.integrality_tolerance); } else { - new_val = val + delta; + new_val = val + mtm_delta; }cpp/src/dual_simplex/dual_simplex_features.hpp (1)
16-16: Add explicit<limits>include forstd::numeric_limits.This header uses
std::numeric_limits(Line 93). Relying on transitive includes is brittle in public headers.♻️ Proposed patch
`#include` <raft/common/nvtx.hpp> `#include` <cstdio> +#include <limits>cpp/src/utilities/hashing.hpp (1)
28-40: Confirm device support forstd::memcpyin the__CUDA_ARCH__path.On some CUDA toolchains,
std::memcpyis not device-available. If nvcc rejects this, switch to byte access viareinterpret_cast(device-safe) orcuda::std::memcpy.🛠️ Device-safe alternative
- uint8_t byte_contents[sizeof(i_t)]; - std::memcpy(byte_contents, &val, sizeof(i_t)); - for (size_t i = 0; i < sizeof(i_t); ++i) { - hash ^= byte_contents[i]; + auto bytes = reinterpret_cast<const uint8_t*>(&val); + for (size_t i = 0; i < sizeof(i_t); ++i) { + hash ^= bytes[i];cpp/src/utilities/models/bounds_strengthening_predictor/quantize.cpp (1)
80-109: Guardfidbounds and avoid hard-coded threshold count.If
fidever exceeds the array size, this will read out of bounds. Also,offset == 188is brittle if thresholds regenerate.🛠️ Proposed hardening
int bounds_strengthening_predictor::quantize(double val, unsigned fid) { + constexpr size_t num_features = sizeof(th_begin) / sizeof(th_begin[0]); + constexpr size_t num_thresholds = sizeof(threshold) / sizeof(threshold[0]); + if (fid >= num_features) { return -10; } const size_t offset = th_begin[fid]; const double* array = &threshold[offset]; int len = th_len[fid]; @@ - if (offset == 188 || val < array[0]) { return -10; } + if (len == 0 || offset == num_thresholds || val < array[0]) { return -10; }cpp/src/utilities/models/pdlp_predictor/quantize.cpp (1)
975-987: Missing bounds check onfidparameter.The function accesses
th_begin[fid]andth_len[fid]without validating thatfidis within bounds. Iffid >= 8, this results in undefined behavior. While the sentinel check (offset == 936) handles unused features, it doesn't protect against invalid feature indices.Consider adding a bounds check or documenting the precondition that
fidmust be less than the number of features:🔒 Suggested defensive check
int pdlp_predictor::quantize(float val, unsigned fid) { + constexpr size_t num_features = sizeof(th_begin) / sizeof(th_begin[0]); + if (fid >= num_features) { return -10; } const size_t offset = th_begin[fid]; const float* array = &threshold[offset];cpp/src/mip/local_search/local_search.cu (1)
193-203: Consider exception safety in cleanup.If
deregister_producerthrows,request_termination()won't be called, potentially leaving the thread running. Consider ensuring cleanup completes even on error:♻️ Suggested improvement
template <typename i_t, typename f_t> void local_search_t<i_t, f_t>::stop_cpufj_deterministic() { if (deterministic_cpu_fj.fj_cpu) { if (deterministic_cpu_fj.fj_cpu->producer_sync) { - deterministic_cpu_fj.fj_cpu->producer_sync->deregister_producer( - &deterministic_cpu_fj.fj_cpu->work_units_elapsed); + try { + deterministic_cpu_fj.fj_cpu->producer_sync->deregister_producer( + &deterministic_cpu_fj.fj_cpu->work_units_elapsed); + } catch (...) { + // Log but continue with termination + } } deterministic_cpu_fj.request_termination(); } }cpp/src/mip/local_search/local_search.cuh (1)
129-132: Consider member visibility.The
deterministic_cpu_fjmember is public alongside other cpu_fj members. This is consistent with the existing pattern (ls_cpu_fj,scratch_cpu_fj,scratch_cpu_fj_on_lp_opt), but direct access to these members from outside the class could lead to race conditions if not carefully managed.cpp/src/utilities/models/fj_predictor/header.h (1)
20-25: Consider using C++ header equivalents.The code uses C-style headers (
float.h,math.h, etc.) instead of their C++ equivalents (<cfloat>,<cmath>,<cstdint>,<cstdio>,<cstdlib>,<cstring>). While this works, C++ headers place symbols in thestdnamespace and are the idiomatic choice for C++ code.The static analysis error about
float.hnot found is likely a false positive (toolchain configuration issue), as this is a standard header.cpp/src/utilities/models/fj_predictor/quantize.cpp (2)
1149-1161: Missing bounds check onfidparameter (consistent with similar issue in pdlp_predictor).Same concern as in
pdlp_predictor::quantize- the function doesn't validate thatfidis within the valid range (0-11 for this predictor with 12 features). Consider adding defensive bounds checking.🔒 Suggested defensive check
int fj_predictor::quantize(double val, unsigned fid) { + constexpr size_t num_features = sizeof(th_begin) / sizeof(th_begin[0]); + if (fid >= num_features) { return -10; } const size_t offset = th_begin[fid]; const double* array = &threshold[offset];
1-19: Note: Blank line at start of file.Line 1 is empty before the license header. This is a minor formatting inconsistency.
cpp/src/dual_simplex/pseudo_costs.hpp (1)
10-19: Add direct includes fordetail::compute_hashandstd::numeric_limits.This header currently relies on transitive includes for
detail::compute_hashandstd::numeric_limits. Making it self‑contained avoids brittle build breaks.♻️ Proposed include fix
`#include` <dual_simplex/types.hpp> `#include` <utilities/omp_helpers.hpp> +#include <utilities/hashing.hpp> `#include` <omp.h> `#include` <cmath> `#include` <cstdint> +#include <limits>Also applies to: 83-84, 139-148
cpp/tests/linear_programming/c_api_tests/c_api_tests.cpp (1)
146-154: Consider adding a test scenario with higherwork_limitfor deeper exploration.All scenarios use
work_limit=2, which limits exploration depth. While this is efficient for CI, consider adding at least one scenario with a higher work limit (e.g., 10-20) to validate determinism during deeper tree exploration where more synchronization events occur.cpp/tests/linear_programming/c_api_tests/c_api_test.c (1)
1291-1295: Consider adding WORK_LIMIT totermination_status_to_stringfor clearer logs.
Right now a work-limit termination will print as “Unknown,” which makes test logs harder to interpret.cpp/src/utilities/models/cpufj_predictor/header.h (1)
8-13: Use C++ headers to avoid C-include portability issues.
Static analysis flagged<float.h>as missing; switching to<cfloat>/<cmath>and friends is more robust in C++ headers.♻️ Suggested include update
-#include <float.h> -#include <math.h> -#include <stdint.h> -#include <stdio.h> -#include <stdlib.h> -#include <string.h> +#include <cfloat> +#include <cmath> +#include <cstdint> +#include <cstdio> +#include <cstdlib> +#include <cstring>cpp/src/utilities/models/cpufj_predictor/quantize.cpp (1)
274-274: Replace magic number231with a named constant.The value
231represents the total number of thresholds and is used to detect unused features. Using a named constant improves readability and maintainability.♻️ Proposed fix
+static constexpr size_t TOTAL_THRESHOLDS = 231; + int cpufj_predictor::quantize(double val, unsigned fid) { const size_t offset = th_begin[fid]; const double* array = &threshold[offset]; int len = th_len[fid]; int low = 0; int high = len; int mid; double mval; - if (offset == 231 || val < array[0]) { return -10; } + if (offset == TOTAL_THRESHOLDS || val < array[0]) { return -10; }cpp/src/utilities/models/dualsimplex_predictor/header.h (1)
8-13: Prefer C++ standard headers over C headers.For C++ code, prefer
<cfloat>,<cmath>,<cstdint>,<cstdio>,<cstdlib>, and<cstring>over their C counterparts. This places declarations in thestdnamespace.♻️ Proposed fix
-#include <float.h> -#include <math.h> -#include <stdint.h> -#include <stdio.h> -#include <stdlib.h> -#include <string.h> +#include <cfloat> +#include <cmath> +#include <cstdint> +#include <cstdio> +#include <cstdlib> +#include <cstring>cpp/src/mip/solver_context.cuh (1)
67-68: Consider making the synchronization interval configurable.The 5.0-second synchronization interval is hardcoded. For tuning deterministic mode performance across different problem sizes and hardware, consider making this configurable through solver settings.
cpp/src/dual_simplex/sparse_vector.cpp (1)
90-119: Consider templating to reduce duplication betweenstd::vectorandins_vectoroverloads.The
to_denseandscatterimplementations forins_vectorare nearly identical to theirstd::vectorcounterparts. A templated helper could eliminate this duplication.♻️ Optional: Template-based approach to reduce duplication
// Private helper template template <typename i_t, typename f_t, typename VectorF> void to_dense_impl(const ins_vector<i_t>& indices, const ins_vector<f_t>& values, i_t n, VectorF& x_dense) { x_dense.clear(); x_dense.resize(n, 0.0); const i_t nz = indices.size(); for (i_t k = 0; k < nz; ++k) { x_dense[indices[k]] = values[k]; } }cpp/src/utilities/work_unit_scheduler.hpp (1)
58-60: Consider makingsync_interval_private.
sync_interval_is declared public (lines 59-60) but there are alreadyset_sync_interval()andget_sync_interval()accessors (lines 39-40). Making the member private would improve encapsulation consistency.♻️ Proposed fix
public: bool verbose{false}; - double sync_interval_; private: + double sync_interval_; double current_sync_target() const;cpp/src/mip/problem/problem.cu (1)
1635-1638: Static variables may cause data races in multi-threaded contexts.
total_time_takenandtotal_callsare static variables that can be concurrently modified by multiple threads. While this is only used for trace logging and doesn't affect correctness, it could produce misleading statistics.Consider using
thread_localif per-thread statistics are acceptable, orstd::atomicif aggregate statistics are needed:- [[maybe_unused]] static double total_time_taken = 0.; - [[maybe_unused]] static int total_calls = 0; + [[maybe_unused]] static thread_local double total_time_taken = 0.; + [[maybe_unused]] static thread_local int total_calls = 0;cpp/src/dual_simplex/bb_event.hpp (1)
10-13: Missing<limits>header.The code uses
std::numeric_limits<f_t>::infinity()inmake_infeasibleandmake_numerical(lines 125, 137), but<limits>is not included. While it may be transitively included, explicit inclusion is safer.Proposed fix
`#include` <algorithm> `#include` <cstdint> +#include <limits> `#include` <tuple> `#include` <vector>cpp/src/dual_simplex/sparse_matrix.hpp (1)
20-21: Consider movingusingdeclaration inside the namespace.The
using cuopt::ins_vector;declaration at global scope (line 21) is outside any namespace. This could pollute the global namespace for all includers. Consider moving it insidecuopt::linear_programming::dual_simplexnamespace for better encapsulation.Proposed fix
-// Import instrumented vector -using cuopt::ins_vector; - namespace cuopt::linear_programming::dual_simplex { +// Import instrumented vector type for this namespace +using cuopt::ins_vector; + template <typename i_t, typename f_t> class csr_matrix_t; // Forward declarationcpp/src/utilities/work_limit_timer.hpp (1)
29-44: Thread safety assumption should be documented.
work_limit_context_thas shared mutable state (global_work_units_elapsed) that can be accessed by multiple threads. The class assumes external synchronization (e.g., BSP barriers). Consider adding a comment documenting this requirement.Suggested documentation
+// Thread safety: This struct is NOT thread-safe. Callers must ensure proper +// synchronization (e.g., BSP barriers) when accessing from multiple threads. +// In deterministic BSP mode, workers synchronize at horizon boundaries. struct work_limit_context_t { double global_work_units_elapsed{0.0};cpp/src/dual_simplex/singletons.cpp (1)
204-213: Fragile static_cast assumes VectorI is always convertible to std::vector<i_t>&.The
static_cast<std::vector<i_t>&>works becauseins_vectorhas an implicit conversion operator toT&, but this is an implicit assumption that may break if other vector-like types are used. The subsequent.underlying().cbegin()calls (lines 208-209) also assume theins_vectorwrapper's interface.Consider using a helper or traits to access the underlying container more explicitly, or document that
VectorImust either bestd::vectororins_vector.cpp/src/utilities/memory_instrumentation.hpp (1)
26-26: Remove debug comment.The comment "Thank you Cursor!" appears to be a development artifact that should be removed before merging.
Proposed fix
- * Example: -* ins_vector<int> vec; // Instrumented std::vector<int> -* vec.push_back(42); -* auto val = vec[0]; -* // When enabled: tracking occurs, counters accumulate -* // When disabled: direct passthrough, compiler optimizes away all overhead -*/ -// Thank you Cursor! + * Example: + * ins_vector<int> vec; // Instrumented std::vector<int> + * vec.push_back(42); + * auto val = vec[0]; + * // When enabled: tracking occurs, counters accumulate + * // When disabled: direct passthrough, compiler optimizes away all overhead + */cpp/src/dual_simplex/branch_and_bound.hpp (1)
357-357: Duplicateprivate:declaration.There's a redundant
private:on line 357 when the class is already in private scope from line 146.cpp/src/dual_simplex/bb_worker_state.hpp (1)
206-216: No bounds validation on variable index.
queue_pseudo_cost_update()directly indexes into snapshot vectors usingvariablewithout bounds checking. Ifvariableis out of range, this causes undefined behavior.Consider adding a debug assertion to catch invalid indices during development:
Proposed fix
void queue_pseudo_cost_update(i_t variable, rounding_direction_t direction, f_t delta) { + assert(variable >= 0 && static_cast<size_t>(variable) < pc_sum_down_snapshot.size()); pseudo_cost_updates.push_back({variable, direction, delta, clock, worker_id}); if (direction == rounding_direction_t::DOWN) {cpp/src/dual_simplex/bsp_debug.hpp (2)
221-243: Silent failure when directory creation fails.In
set_settings(), ifcreate_directories()fails (line 224), all flags are disabled but no warning is logged. Users may not realize their debug output is being silently suppressed.Consider logging a warning to stderr when the directory cannot be created:
Proposed improvement
try { std::filesystem::create_directories(settings_.output_dir); } catch (const std::filesystem::filesystem_error& e) { + std::fprintf(stderr, "Warning: BSP debug directory '%s' could not be created: %s. Debug output disabled.\n", + settings_.output_dir.c_str(), e.what()); settings_.enable_event_log = false; settings_.enable_timeline = false;
253-274: Thread safety assumption:log_horizon_startassumes single-threaded access.
log_horizon_start()modifies shared state (current_horizon_,horizon_wut_start_,timeline_events_) without acquiring the mutex. This is safe if called only by the BSP coordinator during the sync phase when workers are idle, but this assumption should be documented.
| if (num_diving_workers > 0) { | ||
| std::vector<bnb_worker_type_t> diving_types = {bnb_worker_type_t::PSEUDOCOST_DIVING, | ||
| bnb_worker_type_t::LINE_SEARCH_DIVING, | ||
| bnb_worker_type_t::GUIDED_DIVING, | ||
| bnb_worker_type_t::COEFFICIENT_DIVING}; | ||
| bsp_diving_workers_ = std::make_unique<bsp_diving_worker_pool_t<i_t, f_t>>(); | ||
| bsp_diving_workers_->initialize(num_diving_workers, | ||
| diving_types, | ||
| original_lp_, | ||
| Arow, | ||
| var_types_, | ||
| settings_.refactor_frequency, | ||
| settings_.deterministic); | ||
| calculate_variable_locks(original_lp_, var_up_locks_, var_down_locks_); | ||
| } |
There was a problem hiding this comment.
Deterministic BSP ignores per-heuristic enable flags.
BSP always initializes diving workers with all four strategies, even when individual heuristics are disabled in settings. This diverges from the non-BSP path and changes user-configured behavior.
🛠️ Suggested fix (respect settings flags)
- if (num_diving_workers > 0) {
- std::vector<bnb_worker_type_t> diving_types = {bnb_worker_type_t::PSEUDOCOST_DIVING,
- bnb_worker_type_t::LINE_SEARCH_DIVING,
- bnb_worker_type_t::GUIDED_DIVING,
- bnb_worker_type_t::COEFFICIENT_DIVING};
+ if (num_diving_workers > 0) {
+ std::vector<bnb_worker_type_t> diving_types;
+ if (settings_.diving_settings.pseudocost_diving) {
+ diving_types.push_back(bnb_worker_type_t::PSEUDOCOST_DIVING);
+ }
+ if (settings_.diving_settings.line_search_diving) {
+ diving_types.push_back(bnb_worker_type_t::LINE_SEARCH_DIVING);
+ }
+ if (settings_.diving_settings.guided_diving) {
+ diving_types.push_back(bnb_worker_type_t::GUIDED_DIVING);
+ }
+ if (settings_.diving_settings.coefficient_diving) {
+ diving_types.push_back(bnb_worker_type_t::COEFFICIENT_DIVING);
+ }
+ if (diving_types.empty()) return;
bsp_diving_workers_ = std::make_unique<bsp_diving_worker_pool_t<i_t, f_t>>();
bsp_diving_workers_->initialize(num_diving_workers,
diving_types,
original_lp_,
Arow,
var_types_,
settings_.refactor_frequency,
settings_.deterministic);
- calculate_variable_locks(original_lp_, var_up_locks_, var_down_locks_);
+ if (std::find(diving_types.begin(),
+ diving_types.end(),
+ bnb_worker_type_t::COEFFICIENT_DIVING) != diving_types.end()) {
+ calculate_variable_locks(original_lp_, var_up_locks_, var_down_locks_);
+ }
}🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/branch_and_bound.cpp` around lines 1869 - 1883, BSP path
currently unconditionally populates diving_types with all four strategies;
instead, filter by the per-heuristic enable flags in settings before calling
bsp_diving_workers_->initialize. In the block that constructs diving_types (used
by bsp_diving_workers_), check each corresponding settings flag (e.g., the
settings_ flags that control pseudocost, line-search, guided and coefficient
diving) and push only enabled bnb_worker_type_t values; if the resulting vector
is empty, skip creating/initializing bsp_diving_workers_ so BSP remains
consistent with the non-BSP behavior. Ensure you still call
calculate_variable_locks(original_lp_, var_up_locks_, var_down_locks_) only when
workers are actually initialized.
| // Infeasible solutions from GPU heuristics are queued for repair; process them now | ||
| { | ||
| std::vector<std::vector<f_t>> to_repair; | ||
| // TODO: support repair queue in BSP mode | ||
| // mutex_repair_.lock(); | ||
| // if (repair_queue_.size() > 0) { | ||
| // to_repair = repair_queue_; | ||
| // repair_queue_.clear(); | ||
| // } | ||
| // mutex_repair_.unlock(); | ||
|
|
||
| std::sort(to_repair.begin(), | ||
| to_repair.end(), | ||
| [](const std::vector<f_t>& a, const std::vector<f_t>& b) { return a < b; }); | ||
|
|
||
| if (to_repair.size() > 0) { | ||
| settings_.log.debug("BSP sync: Attempting to repair %ld injected solutions\n", | ||
| to_repair.size()); | ||
| for (const std::vector<f_t>& potential_solution : to_repair) { | ||
| std::vector<f_t> repaired_solution; | ||
| f_t repaired_obj; | ||
| bool success = | ||
| repair_solution(edge_norms_, potential_solution, repaired_obj, repaired_solution); | ||
| if (success) { | ||
| // Queue repaired solution with work unit timestamp (...workstamp?) | ||
| mutex_heuristic_queue_.lock(); | ||
| heuristic_solution_queue_.push_back( | ||
| {std::move(repaired_solution), repaired_obj, bsp_current_horizon_}); | ||
| mutex_heuristic_queue_.unlock(); | ||
| } | ||
| } | ||
| } | ||
| } |
There was a problem hiding this comment.
BSP mode drops/never repairs infeasible heuristic solutions.
set_new_solution_deterministic() enqueues infeasible solutions into repair_queue_, but process_history_and_sync() currently skips draining it (TODO is commented). That leaves stale entries and discards potentially repairable incumbents in deterministic runs.
🛠️ Suggested fix (deterministic drain at sync)
- // TODO: support repair queue in BSP mode
- // mutex_repair_.lock();
- // if (repair_queue_.size() > 0) {
- // to_repair = repair_queue_;
- // repair_queue_.clear();
- // }
- // mutex_repair_.unlock();
+ mutex_repair_.lock();
+ if (!repair_queue_.empty()) {
+ to_repair = repair_queue_;
+ repair_queue_.clear();
+ }
+ mutex_repair_.unlock();If you want, I can extend this with deterministic ordering and wut stamping for repaired solutions.
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/branch_and_bound.cpp` around lines 2713 - 2745,
process_history_and_sync() currently ignores repair_queue_ (TODO commented) so
infeasible solutions enqueued by set_new_solution_deterministic() are never
processed; fix by draining repair_queue_ into the local to_repair vector at the
start of the repair block: lock mutex_repair_, move or swap repair_queue_ into
to_repair, clear repair_queue_, unlock mutex_repair_, then proceed with the
existing sort/repair loop and push repaired entries into
heuristic_solution_queue_ (using mutex_heuristic_queue_ and
bsp_current_horizon_); ensure the move preserves deterministic ordering (e.g.,
preserve insertion order before sorting) so deterministic runs actually repair
queued solutions.
| f_t branch_and_bound_t<i_t, f_t>::compute_bsp_lower_bound() | ||
| { | ||
| // Compute lower bound from BFS worker local structures only | ||
| const f_t inf = std::numeric_limits<f_t>::infinity(); | ||
| f_t lower_bound = inf; | ||
|
|
||
| // Check all BFS worker queues | ||
| for (const auto& worker : *bsp_workers_) { | ||
| // Check paused node (current_node) | ||
| if (worker.current_node != nullptr) { | ||
| lower_bound = std::min(worker.current_node->lower_bound, lower_bound); | ||
| } | ||
|
|
||
| // Check plunge stack nodes | ||
| for (auto* node : worker.plunge_stack) { | ||
| lower_bound = std::min(node->lower_bound, lower_bound); | ||
| } | ||
|
|
||
| // Check backlog heap nodes | ||
| for (auto* node : worker.backlog.data()) { | ||
| lower_bound = std::min(node->lower_bound, lower_bound); | ||
| } | ||
| } | ||
|
|
||
| return lower_bound; | ||
| } |
There was a problem hiding this comment.
BSP lower bound ignores lower_bound_ceiling_.
compute_bsp_lower_bound() only scans worker queues. The global lower_bound_ceiling_ (updated for numerical issues) is never factored in, which can yield inf and distort gap checks/termination.
🛠️ Suggested fix
- const f_t inf = std::numeric_limits<f_t>::infinity();
- f_t lower_bound = inf;
+ const f_t inf = std::numeric_limits<f_t>::infinity();
+ f_t lower_bound = lower_bound_ceiling_.load();
+ if (!std::isfinite(lower_bound)) lower_bound = inf;
...
- return lower_bound;
+ return std::min(lower_bound_ceiling_.load(), lower_bound);📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| f_t branch_and_bound_t<i_t, f_t>::compute_bsp_lower_bound() | |
| { | |
| // Compute lower bound from BFS worker local structures only | |
| const f_t inf = std::numeric_limits<f_t>::infinity(); | |
| f_t lower_bound = inf; | |
| // Check all BFS worker queues | |
| for (const auto& worker : *bsp_workers_) { | |
| // Check paused node (current_node) | |
| if (worker.current_node != nullptr) { | |
| lower_bound = std::min(worker.current_node->lower_bound, lower_bound); | |
| } | |
| // Check plunge stack nodes | |
| for (auto* node : worker.plunge_stack) { | |
| lower_bound = std::min(node->lower_bound, lower_bound); | |
| } | |
| // Check backlog heap nodes | |
| for (auto* node : worker.backlog.data()) { | |
| lower_bound = std::min(node->lower_bound, lower_bound); | |
| } | |
| } | |
| return lower_bound; | |
| } | |
| f_t branch_and_bound_t<i_t, f_t>::compute_bsp_lower_bound() | |
| { | |
| // Compute lower bound from BFS worker local structures only | |
| const f_t inf = std::numeric_limits<f_t>::infinity(); | |
| f_t lower_bound = lower_bound_ceiling_.load(); | |
| if (!std::isfinite(lower_bound)) lower_bound = inf; | |
| // Check all BFS worker queues | |
| for (const auto& worker : *bsp_workers_) { | |
| // Check paused node (current_node) | |
| if (worker.current_node != nullptr) { | |
| lower_bound = std::min(worker.current_node->lower_bound, lower_bound); | |
| } | |
| // Check plunge stack nodes | |
| for (auto* node : worker.plunge_stack) { | |
| lower_bound = std::min(node->lower_bound, lower_bound); | |
| } | |
| // Check backlog heap nodes | |
| for (auto* node : worker.backlog.data()) { | |
| lower_bound = std::min(node->lower_bound, lower_bound); | |
| } | |
| } | |
| return std::min(lower_bound_ceiling_.load(), lower_bound); | |
| } |
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/branch_and_bound.cpp` around lines 3041 - 3066,
compute_bsp_lower_bound() only reads per-worker nodes and can return inf because
it omits the global lower_bound_ceiling_; include lower_bound_ceiling_ in the
computed minimum so numerical ceilings are enforced. Update the function
(branch_and_bound_t<i_t, f_t>::compute_bsp_lower_bound) to initialize
lower_bound to std::min(inf, lower_bound_ceiling_) or explicitly take
lower_bound = lower_bound_ceiling_ and then min with
worker.current_node->lower_bound, worker.plunge_stack nodes, and
worker.backlog.data() nodes before returning.
| manifold.add("A_transpose.col_start", A_transpose.col_start); | ||
| manifold.add("A_transpose.i", A_transpose.i); | ||
| manifold.add("A_transpose.x", A_transpose.x); | ||
| { | ||
| PHASE2_NVTX_RANGE("DualSimplex::transpose_A"); | ||
| lp.A.transpose(A_transpose); | ||
| } |
There was a problem hiding this comment.
A_transpose is registered in the manifold twice.
manifold.add("A_transpose.*", …) occurs before transpose and then again later. If instrumentation_manifold_t::add() expects unique keys, this can overwrite or double-count loads/stores and skew work-unit prediction.
🛠️ Suggested fix (keep a single registration)
- // Add A_transpose matrix arrays to manifold for memory tracking
- manifold.add("A_transpose.col_start", A_transpose.col_start);
- manifold.add("A_transpose.i", A_transpose.i);
- manifold.add("A_transpose.x", A_transpose.x);Also applies to: 2557-2560
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/phase2.cpp` around lines 2486 - 2492, The code registers
A_transpose into the instrumentation manifold twice via calls to
manifold.add("A_transpose.col_start"/"A_transpose.i"/"A_transpose.x") both
before and after lp.A.transpose(A_transpose); remove the duplicate registration
so each A_transpose.* key is added only once (choose whether to register before
or after transpose and delete the other set), and apply the same
single-registration fix to the other duplicate group around the second transpose
(the block at the later occurrence). Ensure only one set of manifold.add calls
references A_transpose.col_start, A_transpose.i, and A_transpose.x.
|
|
||
| CUOPT_LOG_DEBUG("Determinism mode: %s", | ||
| context.settings.determinism_mode == CUOPT_MODE_DETERMINISTIC ? "deterministic" | ||
| : "opportunistic"); | ||
|
|
||
| // Debug: Allow disabling GPU heuristics to test B&B tree determinism in isolation | ||
| const char* disable_heuristics_env = std::getenv("CUOPT_DISABLE_GPU_HEURISTICS"); | ||
| if (context.settings.determinism_mode == CUOPT_MODE_DETERMINISTIC) { | ||
| CUOPT_LOG_INFO("Running deterministic mode with CPUFJ heuristic"); | ||
| population.initialize_population(); | ||
| population.allocate_solutions(); | ||
|
|
||
| // Start CPUFJ in deterministic mode with B&B integration | ||
| if (context.branch_and_bound_ptr != nullptr) { | ||
| ls.start_cpufj_deterministic(*context.branch_and_bound_ptr); | ||
| } | ||
|
|
||
| while (!check_b_b_preemption()) { | ||
| std::this_thread::sleep_for(std::chrono::milliseconds(100)); | ||
| } | ||
|
|
||
| // Stop CPUFJ when B&B is done | ||
| ls.stop_cpufj_deterministic(); | ||
|
|
||
| population.add_external_solutions_to_population(); | ||
| return population.best_feasible(); | ||
| } | ||
| if (disable_heuristics_env != nullptr && std::string(disable_heuristics_env) == "1") { | ||
| CUOPT_LOG_INFO("GPU heuristics disabled via CUOPT_DISABLE_GPU_HEURISTICS=1"); | ||
| population.initialize_population(); | ||
| population.allocate_solutions(); | ||
|
|
||
| while (!check_b_b_preemption()) { | ||
| std::this_thread::sleep_for(std::chrono::milliseconds(100)); | ||
| } | ||
| return population.best_feasible(); | ||
| } |
There was a problem hiding this comment.
Potential infinite wait when B&B preemption never arrives.
Both branches spin until check_b_b_preemption() is true, but they ignore time_limit and can block forever if B&B isn’t running or never signals preemption. Add a timeout/guard (and a null‑B&B fallback) to prevent hangs.
🛠️ Proposed fix (timeout + null‑B&B guard)
- // Start CPUFJ in deterministic mode with B&B integration
- if (context.branch_and_bound_ptr != nullptr) {
- ls.start_cpufj_deterministic(*context.branch_and_bound_ptr);
- }
+ // Start CPUFJ in deterministic mode with B&B integration
+ if (context.branch_and_bound_ptr != nullptr) {
+ ls.start_cpufj_deterministic(*context.branch_and_bound_ptr);
+ } else {
+ CUOPT_LOG_WARN("Deterministic mode requested without B&B; returning current best.");
+ return population.best_feasible();
+ }
- while (!check_b_b_preemption()) {
- std::this_thread::sleep_for(std::chrono::milliseconds(100));
- }
+ while (!check_b_b_preemption()) {
+ if (timer.check_time_limit()) {
+ CUOPT_LOG_WARN("Timed out waiting for B&B preemption in deterministic mode");
+ break;
+ }
+ std::this_thread::sleep_for(std::chrono::milliseconds(100));
+ }
@@
- while (!check_b_b_preemption()) {
- std::this_thread::sleep_for(std::chrono::milliseconds(100));
- }
+ while (!check_b_b_preemption()) {
+ if (timer.check_time_limit()) {
+ CUOPT_LOG_WARN("Timed out waiting for B&B preemption (GPU heuristics disabled)");
+ break;
+ }
+ std::this_thread::sleep_for(std::chrono::milliseconds(100));
+ }🤖 Prompt for AI Agents
In `@cpp/src/mip/diversity/diversity_manager.cu` around lines 309 - 345, The
current deterministic and "heuristics disabled" branches spin on
check_b_b_preemption() with no timeout and no null-B&B fallback, which can block
forever; update both loops to enforce a maximum wait (e.g., derive from a
setting or use a reasonable default) and break if the timeout elapses, call
ls.stop_cpufj_deterministic() if CPUFJ was started, and if
context.branch_and_bound_ptr is null skip waiting entirely and return
population.best_feasible() immediately; make these changes around the existing
symbols population.initialize_population(), population.allocate_solutions(),
ls.start_cpufj_deterministic(), ls.stop_cpufj_deterministic(), and
population.best_feasible() so the code never hangs waiting for B&B.
| typename model_t::Entry data[model_t::NUM_FEATURES]; | ||
| for (int i = 0; i < model_t::NUM_FEATURES; ++i) { | ||
| if (features.find(std::string(model_t::feature_names[i])) == features.end()) { | ||
| data[i].missing = -1; | ||
| CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]); | ||
| } else { | ||
| data[i].fvalue = features.at(std::string(model_t::feature_names[i])); | ||
| } | ||
| } |
There was a problem hiding this comment.
Initialize Entry::missing for non‑missing features.
data[i].missing is left uninitialized in the “present feature” path, which can misclassify features and yield nondeterministic cache keys/predictions.
🐛 Proposed fix
- for (int i = 0; i < model_t::NUM_FEATURES; ++i) {
- if (features.find(std::string(model_t::feature_names[i])) == features.end()) {
- data[i].missing = -1;
- CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]);
- } else {
- data[i].fvalue = features.at(std::string(model_t::feature_names[i]));
- }
- }
+ for (int i = 0; i < model_t::NUM_FEATURES; ++i) {
+ auto it = features.find(std::string(model_t::feature_names[i]));
+ if (it == features.end()) {
+ data[i].missing = -1;
+ data[i].fvalue = 0.0f;
+ CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]);
+ } else {
+ data[i].missing = 0;
+ data[i].fvalue = it->second;
+ }
+ }📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| typename model_t::Entry data[model_t::NUM_FEATURES]; | |
| for (int i = 0; i < model_t::NUM_FEATURES; ++i) { | |
| if (features.find(std::string(model_t::feature_names[i])) == features.end()) { | |
| data[i].missing = -1; | |
| CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]); | |
| } else { | |
| data[i].fvalue = features.at(std::string(model_t::feature_names[i])); | |
| } | |
| } | |
| typename model_t::Entry data[model_t::NUM_FEATURES]; | |
| for (int i = 0; i < model_t::NUM_FEATURES; ++i) { | |
| auto it = features.find(std::string(model_t::feature_names[i])); | |
| if (it == features.end()) { | |
| data[i].missing = -1; | |
| data[i].fvalue = 0.0f; | |
| CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]); | |
| } else { | |
| data[i].missing = 0; | |
| data[i].fvalue = it->second; | |
| } | |
| } |
🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_predictor.cpp` around lines 44 - 52, The loop
currently leaves model_t::Entry::missing uninitialized for present features
causing nondeterministic behavior; in the else branch where you set
data[i].fvalue (and when features.at(...) succeeds), also explicitly set
data[i].missing = 0 (and initialize any other Entry fields that must have
deterministic defaults) so every data[i] has a defined missing flag; update the
block that references model_t::feature_names, features.at(...), and
CUOPT_LOG_WARN to set missing=0 in the present-feature path.
There was a problem hiding this comment.
this is a union
There was a problem hiding this comment.
@aliceb-nv, you're absolutely right. Since Entry is a union, the current code is correct.
In a union, all fields share the same memory location, so when you set data[i].fvalue for present features, the memory is initialized—even though missing would contain the reinterpreted bit pattern of the float value. The pattern used here is standard for discriminated unions:
- Feature missing: Set
data[i].missing = -1(the discriminant value) - Feature present: Set
data[i].fvalue = features.at(...)(the actual value)
The model's predict function likely checks if missing == -1 to determine whether to treat the entry as missing or to use the fvalue. The current implementation is correct and doesn't need the additional initialization that CodeRabbit suggested.
🧠 Learnings used
Learnt from: aliceb-nv
Repo: NVIDIA/cuopt PR: 798
File: cpp/src/utilities/models/fj_predictor/quantize.cpp:1149-1161
Timestamp: 2026-01-26T14:17:20.839Z
Learning: The files in cpp/src/utilities/models/*/quantize.cpp (fj_predictor, cpufj_predictor, pdlp_predictor, dualsimplex_predictor, bounds_strengthening_predictor) are autogenerated model artifacts and should not have manual code changes suggested; any improvements should target the model generation tooling or calling code validation layer.
Learnt from: aliceb-nv
Repo: NVIDIA/cuopt PR: 798
File: cpp/src/utilities/models/pdlp_predictor/quantize.cpp:975-987
Timestamp: 2026-01-26T14:17:22.967Z
Learning: Files in cpp/src/utilities/models/*/quantize.cpp (and related model predictor files like header.h, main.cpp) are autogenerated and should not receive suggestions for manual code changes. These are ML model artifacts (e.g., fj_predictor, cpufj_predictor, dualsimplex_predictor, pdlp_predictor, bounds_strengthening_predictor).
Learnt from: CR
Repo: NVIDIA/cuopt PR: 0
File: .github/.coderabbit_review_guide.md:0-0
Timestamp: 2025-11-25T10:20:49.822Z
Learning: Applies to **/*.{cu,cuh,cpp,hpp,h} : Refactor code duplication in solver components (3+ occurrences) into shared utilities; for GPU kernels, use templated device functions to avoid duplication
| auto cached_it = prediction_cache.find(key); | ||
| if (cached_it != prediction_cache.end()) { return cached_it->second; } | ||
|
|
||
| double result = 0.0; | ||
| auto start = std::chrono::high_resolution_clock::now(); | ||
| model_t::predict(data, 0, &result); | ||
| auto end = std::chrono::high_resolution_clock::now(); | ||
| std::chrono::duration<double, std::milli> elapsed = end - start; | ||
| if (debug) CUOPT_LOG_DEBUG("Prediction time: %f ms", elapsed.count()); | ||
|
|
||
| float scaled_result = scaler_.scale_work_units(result); | ||
| prediction_cache[key] = scaled_result; | ||
| if (debug) CUOPT_LOG_DEBUG("Result: %f (scaled: %f)", result, scaled_result); |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
#!/bin/bash
# Search for work_unit_predictor usage patterns
rg "work_unit_predictor" -g '*.{cpp,hpp,h}' -C 2 | head -100Repository: NVIDIA/cuopt
Length of output: 3983
🏁 Script executed:
#!/bin/bash
# Look for the class definition and cache member
rg -A 30 "class work_unit_predictor" -g '*.{hpp,h}'Repository: NVIDIA/cuopt
Length of output: 868
🏁 Script executed:
#!/bin/bash
# Check for thread-related usage (mutex, locks, atomic operations on cache)
rg "prediction_cache|std::mutex|std::lock|atomic" cpp/src/utilities/work_unit_predictor.cpp -C 1Repository: NVIDIA/cuopt
Length of output: 361
🏁 Script executed:
#!/bin/bash
# Find the function containing the static predictor in branch_and_bound.cpp
rg -B 10 'static.*work_unit_predictor_t<bounds_strengthening_predictor' cpp/src/dual_simplex/branch_and_bound.cpp | head -40Repository: NVIDIA/cuopt
Length of output: 529
🏁 Script executed:
#!/bin/bash
# Check for all static instantiations of work_unit_predictor
rg 'static.*work_unit_predictor_t' -g '*.cpp' -C 5Repository: NVIDIA/cuopt
Length of output: 1011
🏁 Script executed:
#!/bool
# Check if the containing function is called from parallel regions
rg 'bounds_strengthening_predictor|#pragma omp' cpp/src/dual_simplex/branch_and_bound.cpp -C 2 | head -50Repository: NVIDIA/cuopt
Length of output: 1649
Add synchronization to prediction_cache in the static predictor instance.
The static predictor instance in cpp/src/dual_simplex/branch_and_bound.cpp is shared across multiple OpenMP threads (called from exploration_ramp_up() which executes within #pragma omp task and #pragma omp parallel regions). The prediction_cache is accessed and mutated without synchronization, creating a data race. Protect cache operations with a mutex or use thread-safe container. Note: The thread_local instance in cpp/src/dual_simplex/phase2.cpp is already safe.
🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_predictor.cpp` around lines 62 - 74,
prediction_cache is accessed concurrently by the static predictor instance
causing a data race; add a mutex (e.g., std::mutex prediction_cache_mutex as a
member of the predictor class) and protect all accesses to prediction_cache with
a lock (use std::lock_guard<std::mutex> guard(prediction_cache_mutex)) around
the find/return path and around the code that assigns prediction_cache[key] =
scaled_result; keep model_t::predict and scaler_.scale_work_units calls outside
the locked region if possible (or document if they must be inside), ensuring
both lookup and insertion are synchronized to eliminate the race.
| cpu_work_unit_scaler_t() | ||
| { | ||
| constexpr double baseline_max_clock = 3800.0; | ||
| double max_clock = get_cpu_max_clock_mhz(); | ||
| scaling_factor_ = baseline_max_clock / max_clock; | ||
| } |
There was a problem hiding this comment.
Guard against zero/invalid CPU clock readings.
If get_cpu_max_clock_mhz() returns 0 (or invalid), the scaling factor becomes inf and propagates into work limits.
🐛 Proposed guard
cpu_work_unit_scaler_t()
{
constexpr double baseline_max_clock = 3800.0;
double max_clock = get_cpu_max_clock_mhz();
- scaling_factor_ = baseline_max_clock / max_clock;
+ scaling_factor_ = (max_clock > 0.0) ? (baseline_max_clock / max_clock)
+ : 1.0;
}📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| cpu_work_unit_scaler_t() | |
| { | |
| constexpr double baseline_max_clock = 3800.0; | |
| double max_clock = get_cpu_max_clock_mhz(); | |
| scaling_factor_ = baseline_max_clock / max_clock; | |
| } | |
| cpu_work_unit_scaler_t() | |
| { | |
| constexpr double baseline_max_clock = 3800.0; | |
| double max_clock = get_cpu_max_clock_mhz(); | |
| scaling_factor_ = (max_clock > 0.0) ? (baseline_max_clock / max_clock) | |
| : 1.0; | |
| } |
🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_predictor.hpp` around lines 33 - 38, The
cpu_work_unit_scaler_t constructor uses get_cpu_max_clock_mhz() to compute
scaling_factor_, but if max_clock is 0 or negative the result becomes INF;
update the constructor (cpu_work_unit_scaler_t) to guard against invalid
readings by checking if max_clock <= 0 and in that case set a safe fallback
(e.g., set scaling_factor_ = 1.0 or use baseline_max_clock as the divisor) and
optionally log the anomaly; ensure scaling_factor_ is always finite before
storing it.
| private: | ||
| mutable std::unordered_map<uint32_t, float> prediction_cache; | ||
| scaler_t scaler_; |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
fd -t f "work_unit_predictor\.(hpp|cpp)" --exec wc -l {} +Repository: NVIDIA/cuopt
Length of output: 258
🏁 Script executed:
find . -name "work_unit_predictor.hpp" -o -name "work_unit_predictor.cpp" | head -20Repository: NVIDIA/cuopt
Length of output: 143
🏁 Script executed:
cat -n cpp/src/utilities/work_unit_predictor.hppRepository: NVIDIA/cuopt
Length of output: 2441
🏁 Script executed:
cat -n cpp/src/utilities/work_unit_predictor.cppRepository: NVIDIA/cuopt
Length of output: 3947
🏁 Script executed:
rg "work_unit_predictor" --type cpp --type hpp -B 2 -A 2Repository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
rg "work_unit_predictor_t" --type cpp --type hpp | head -40Repository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
rg "work_unit_predictor" --type-list | grep -i hppRepository: NVIDIA/cuopt
Length of output: 445
🏁 Script executed:
rg "work_unit_predictor" -g "*.hpp" -g "*.cpp" -g "*.cu"Repository: NVIDIA/cuopt
Length of output: 1699
🏁 Script executed:
rg -B 3 -A 3 "static cuopt::work_unit_predictor_t" cpp/src/dual_simplex/branch_and_bound.cppRepository: NVIDIA/cuopt
Length of output: 300
🏁 Script executed:
rg -B 5 -A 5 "thread_local.*work_unit_predictor_t" cpp/src/dual_simplex/phase2.cppRepository: NVIDIA/cuopt
Length of output: 625
🏁 Script executed:
rg -B 2 -A 2 "work_unit_predictors" -g "*.cu" -g "*.cpp"Repository: NVIDIA/cuopt
Length of output: 403
🏁 Script executed:
cat -n cpp/src/dual_simplex/branch_and_bound.cpp | sed -n '1,100p'Repository: NVIDIA/cuopt
Length of output: 3879
🏁 Script executed:
rg -B 10 "static cuopt::work_unit_predictor_t" cpp/src/dual_simplex/branch_and_bound.cppRepository: NVIDIA/cuopt
Length of output: 529
🏁 Script executed:
rg "work_unit_predictors" -B 5 -A 5 cpp/src/mip/feasibility_jump/fj_cpu.cuRepository: NVIDIA/cuopt
Length of output: 565
🏁 Script executed:
rg -B 20 "static cuopt::work_unit_predictor_t" cpp/src/dual_simplex/branch_and_bound.cpp | head -40Repository: NVIDIA/cuopt
Length of output: 912
🏁 Script executed:
grep -n "def.*bounds_strengthening\|void.*bounds_strengthening\|auto.*bounds_strengthening" cpp/src/dual_simplex/branch_and_bound.cppRepository: NVIDIA/cuopt
Length of output: 38
🏁 Script executed:
rg "struct.*context|class.*context" cpp/src/mip/feasibility_jump/ -A 5Repository: NVIDIA/cuopt
Length of output: 38
🏁 Script executed:
cat -n cpp/src/dual_simplex/branch_and_bound.cpp | sed -n '500,600p'Repository: NVIDIA/cuopt
Length of output: 5089
🏁 Script executed:
rg "#pragma omp" cpp/src/dual_simplex/branch_and_bound.cppRepository: NVIDIA/cuopt
Length of output: 307
🏁 Script executed:
rg "work_unit_predictors" cpp/src/mip/feasibility_jump -B 10 -A 2Repository: NVIDIA/cuopt
Length of output: 1136
Make prediction_cache thread‑safe.
std::unordered_map isn't safe for concurrent reads/writes. The predictor instance in branch_and_bound.cpp is declared as static and called from OpenMP parallel regions, creating a data race when multiple threads access predict_scalar() simultaneously.
Add a mutable std::mutex to guard cache access:
-#include <unordered_map>
+#include <unordered_map>
+#include <mutex>
private:
mutable std::unordered_map<uint32_t, float> prediction_cache;
+ mutable std::mutex prediction_cache_mutex;
scaler_t scaler_;Lock around all cache operations in predict_scalar() (lines 62–63 and 73 in the .cpp file).
🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_predictor.hpp` around lines 58 - 60, The
prediction_cache is not thread-safe; add a mutable std::mutex member (e.g.,
prediction_cache_mutex) and `#include` <mutex>, then protect all accesses to
prediction_cache inside predict_scalar() with a lock (use
std::lock_guard<std::mutex> or similar) around the find/read and insert/update
paths so concurrent calls from OpenMP threads cannot race while still leaving
scaler_ untouched.
| // Ensure seed is positive int32_t | ||
| auto seed = std::random_device{}() & 0x7fffffff; | ||
| std::cout << "Tested with seed " << seed << "\n"; | ||
| settings.seed = seed; |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
cat -n cpp/tests/mip/determinism_test.cu | head -180Repository: NVIDIA/cuopt
Length of output: 7868
🏁 Script executed:
rg -n 'seed_generator::set_seed|settings\.seed' --type cpp --type-add 'cpp:*.{cpp,cu,hpp,cuh}' -C 2 | head -150Repository: NVIDIA/cuopt
Length of output: 7090
🏁 Script executed:
rg -n 'struct.*settings|seed.*=' cpp/src/linear_programming/mip/solver_settings.hpp cpp/include/cuopt/linear_programming/mip/solver_settings.hpp | head -50Repository: NVIDIA/cuopt
Length of output: 245
🏁 Script executed:
sed -n '160,180p' cpp/src/mip/solve.cuRepository: NVIDIA/cuopt
Length of output: 972
🏁 Script executed:
rg -n 'seed_generator::get_seed|seed_generator::set_seed' cpp/src --type cpp --type-add 'cpp:*.{cpp,cu,hpp,cuh}' -B 1 -A 1Repository: NVIDIA/cuopt
Length of output: 15842
🏁 Script executed:
sed -n '50,142p' cpp/tests/mip/determinism_test.cu | grep -n 'seed_generator\|settings.seed'Repository: NVIDIA/cuopt
Length of output: 109
🏁 Script executed:
grep -n 'settings.seed' cpp/src/mip/feasibility_jump/fj_cpu.cu cpp/src/mip/feasibility_jump/feasibility_jump.cu | head -20Repository: NVIDIA/cuopt
Length of output: 1107
🏁 Script executed:
grep -A 5 "if (settings.seed >= 0)" cpp/src/mip/solve.cuRepository: NVIDIA/cuopt
Length of output: 336
🏁 Script executed:
grep -B 5 -A 5 'settings.seed = cuopt::seed_generator::get_seed' cpp/src/mip/feasibility_jump/fj_cpu.cu cpp/src/mip/feasibility_jump/feasibility_jump.cuRepository: NVIDIA/cuopt
Length of output: 2164
🏁 Script executed:
sed -n '290,300p' cpp/src/dual_simplex/phase2.cppRepository: NVIDIA/cuopt
Length of output: 321
🏁 Script executed:
cat -n cpp/include/cuopt/linear_programming/mip/solver_settings.hpp | grep -A 5 -B 5 'seed'Repository: NVIDIA/cuopt
Length of output: 688
🏁 Script executed:
cat -n cpp/tests/mip/determinism_test.cu | sed -n '124,142p'Repository: NVIDIA/cuopt
Length of output: 959
Fix parameterized test seed handling to match non-parameterized tests.
The parameterized test (DeterministicBBInstanceTest, lines 169-174) calls cuopt::seed_generator::set_seed(seed) but does not set settings.seed. This causes a bug: when solve_mip() is called, settings.seed remains at its default value of -1, so the solver does not initialize the seed generator in solve.cu:171. Additionally, components like dual_simplex/phase2.cpp and feasibility_jump use settings.seed directly to initialize their random number generators, receiving -1 instead of the intended seed value.
Set settings.seed = seed; before calling solve_mip() to ensure consistent seeding across all tests, matching the approach used in non-parameterized tests (lines 71, 104).
🤖 Prompt for AI Agents
In `@cpp/tests/mip/determinism_test.cu` around lines 68 - 71, The parameterized
test (DeterministicBBInstanceTest) calls cuopt::seed_generator::set_seed(seed)
but fails to set settings.seed, so downstream code using settings.seed sees -1;
before invoking solve_mip() in that test, assign settings.seed = seed to match
non-parameterized tests and ensure all RNG initializations (e.g., those reading
settings.seed and the solver's seed setup) receive the intended seed value.
|
/ok to test 0b0e65b |
There was a problem hiding this comment.
Actionable comments posted: 19
Caution
Some comments are outside the diff and can’t be posted inline due to platform limitations.
⚠️ Outside diff range comments (3)
cpp/include/cuopt/linear_programming/mip/solver_settings.hpp (1)
82-111: Document that seed = -1 does not auto-generate a random seed and clarify determinism requirements.Header documentation misleadingly claims "Use -1 to generate a random seed," but code at
cpp/src/mip/solve.cu:171only callsset_seed()whensettings.seed >= 0, leaving seed_generator's static state unchanged (initialized to 0) when seed = -1. Whendeterminism_mode == CUOPT_MODE_DETERMINISTICwith defaultseed = -1, the determinism contract is silently violated. Update the seed parameter documentation to clarify that -1 does not trigger random seed generation, and either add validation to prevent deterministic mode with seed = -1, or auto-override seed to a generated value when determinism is required.cpp/src/dual_simplex/basis_solves.cpp (2)
74-101:fclosecalled unconditionally whenfopenmay fail.If
fopenreturnsNULL, callingfclose(file)at line 100 is undefined behavior. Thefcloseshould be inside theif (file != NULL)block.🐛 Proposed fix
if (file != NULL) { fprintf(file, "m = %d; col_singletons = %d; row_singletons = %d;\n", m, col_singletons, row_singletons); B.print_matrix(file); fprintf(file, "B = sparse(ijx(:,1), ijx(:,2), ijx(:,3), m, m);\n"); fprintf(file, "row_perm = [\n"); for (i_t i = 0; i < m; ++i) { fprintf(file, "%d\n", row_perm[i] + 1); } fprintf(file, "];\n"); fprintf(file, "col_perm=[\n"); for (i_t i = 0; i < m; ++i) { fprintf(file, "%d\n", col_perm[i] + 1); } fprintf(file, "];\n;"); fprintf(file, "row_perm_inv = [\n"); for (i_t i = 0; i < m; ++i) { fprintf(file, "%d\n", row_perm_inv[i] + 1); } fprintf(file, "];\n"); + fclose(file); } - fclose(file); }
116-145: Samefcloseissue inwrite_factor_info.The
fclose(file)at line 144 is outside the null check, causing undefined behavior iffopenfails.🐛 Proposed fix
if (file != NULL) { // ... all fprintf calls ... fprintf(file, "];\n"); + fclose(file); } - fclose(file); }
🤖 Fix all issues with AI agents
In `@cpp/src/dual_simplex/bb_event.hpp`:
- Around line 10-13: The header uses std::numeric_limits in make_infeasible and
make_numerical but does not include <limits>, causing compilation failures; add
the missing `#include` <limits> to cpp/src/dual_simplex/bb_event.hpp (near the
other standard includes) so that std::numeric_limits references in the
make_infeasible and make_numerical implementations resolve correctly.
In `@cpp/src/dual_simplex/bb_worker_state.hpp`:
- Around line 20-27: The counter next_creation_seq in bb_worker_state.hpp can
overflow as an int32_t; change its type to a 64-bit integer (e.g., int64_t or
uint64_t) wherever it is declared and used (including the other related counters
referenced around lines 114-116) to ensure unique, monotonic IDs for large B&B
trees; update any constructors, accessors, comparisons, and serialization code
that reference next_creation_seq (and the other identified sequence/counter
members) to use the new 64-bit type so behavior and ordering remain stable.
- Around line 513-527: The initialize method must guard against an empty
diving_types vector to avoid division-by-zero when computing i %
diving_types.size(); add an early check in initialize that validates
!diving_types.empty() and if empty either throw a clear std::invalid_argument
(e.g., "initialize: diving_types must not be empty") or set a documented default
bnb_worker_type_t before the loop; ensure the check/run uses the symbols
initialize, diving_types, workers_, and bnb_worker_type_t so the fix is applied
right before the for-loop that currently does i % diving_types.size().
- Around line 173-186: reset_for_horizon currently updates events.horizon_start
and events.horizon_end but not the worker state's member variables horizon_start
and horizon_end, leaving them stale until set_snapshots() runs; modify
reset_for_horizon to assign the incoming horizon_start and horizon_end
parameters to the class members (horizon_start and horizon_end) in addition to
setting events.horizon_start/events.horizon_end so state is synchronized
immediately (keep other resets like event_sequence, clock, and work_context
changes intact).
In `@cpp/src/dual_simplex/branch_and_bound.cpp`:
- Around line 1968-1984: total_thread_count currently adds num_diving_workers
even when bsp_diving_workers_ may be null, causing a null deref in the else
branch; compute the actual diving count from bsp_diving_workers_ (e.g.,
actual_diving_count = bsp_diving_workers_ ? bsp_diving_workers_->size() : 0),
use total_thread_count = num_bfs_workers + actual_diving_count, and in the
parallel region guard the diving path by checking diving_id <
actual_diving_count (or only spawn diving threads when actual_diving_count > 0)
before accessing bsp_diving_workers_ and calling run_diving_worker_loop; update
references to thread_id, num_bfs_workers, diving_id, bsp_diving_workers_,
run_diving_worker_loop and total_thread_count accordingly.
In `@cpp/src/dual_simplex/bsp_debug.hpp`:
- Around line 253-274: log_horizon_start and several adjacent methods write to
the shared trace stringstreams without taking the mutex while log_horizon_hash
does, causing inconsistent synchronization; fix by acquiring
std::lock_guard<std::mutex> lock(mutex_) at the start of every method that
mutates trace_ss_ or timeline_events_ (e.g., log_horizon_start,
log_sync_phase_end, log_heuristic_received, log_incumbent_update and any other
trace-writing worker/manager log_* methods) so all writes use the same mutex_;
ensure the lock surrounds trace_ss_ writes and
timeline_events_.clear()/mutations to guarantee consistent thread-safety.
In `@cpp/src/dual_simplex/diving_heuristics.hpp`:
- Around line 114-120: The code computes down_dist and up_dist by accessing
incumbent[j] without bounds checking; mirror the defensive pattern used in
pseudocost_diving_from_arrays by using a conditional fallback when j is out of
range (e.g., use solution[j] if j >= incumbent.size()). Update the loop in
diving_heuristics.hpp where fractional is iterated (variables: solution,
incumbent, fractional, f_down, f_up, down_dist, up_dist, rounding_direction_t,
eps) to compute a safe root_val or incumbent_val with the ternary check before
computing distances, then use that value in down_dist and up_dist calculations
so the rounding_direction_t decision is robust.
In `@cpp/src/mip/feasibility_jump/fj_cpu.cu`:
- Around line 396-401: The bandwidth calculation can divide by zero when
time_window_ms is 0; modify the computation around mem_bandwidth_gb_per_sec to
guard the denominator by checking time_window_ms (used to compute
mem_bandwidth_gb_per_sec) and either skip/assign a sentinel (e.g., 0.0) when
time_window_ms <= 0 or clamp it to a small epsilon before dividing; update the
block that computes mem_loads_mb, mem_stores_mb, mem_total_mb and
mem_bandwidth_gb_per_sec so mem_bandwidth_gb_per_sec is only computed with a
safe non‑zero window.
- Around line 284-341: precompute_problem_features currently divides by n_vars
and n_cstrs producing NaN for zero-dimension models; update
precompute_problem_features to guard all divisions and statistics by checking
n_vars>0 and n_cstrs>0 (use 0/defaults when counts are zero). Specifically,
ensure avg_var_degree, var_degree_cv, var_degree_std, max_var_degree,
var_degrees population and the variance loop are skipped or set to 0 when
n_vars==0, and likewise avg_cstr_degree, cstr_degree_cv, cstr_degree_std,
max_cstr_degree, cstr_degrees and its variance loop are skipped or set to 0 when
n_cstrs==0; finally compute problem_density only when both n_vars>0 and
n_cstrs>0 (otherwise set to 0). Target the function precompute_problem_features
and fields on fj_cpu (avg_var_degree, var_degree_cv, avg_cstr_degree,
cstr_degree_cv, problem_density, max_var_degree, max_cstr_degree) when applying
these guards.
In `@cpp/src/mip/problem/problem_helpers.cuh`:
- Around line 359-361: The early return on zero nnz currently bypasses CSR
validation; move or invoke check_csr_representation (or the CSR validation call
present in this translation unit) before returning when values.size() == 0 so
malformed offsets/indices are caught; specifically, in the function containing
the if (values.size() == 0) { return; } check, call
check_csr_representation(offsets, indices, values) (or the existing validation
helper that inspects offsets/indices) and only return after that validation
succeeds.
In `@cpp/src/mip/problem/problem.cu`:
- Around line 881-927: In compute_auxiliary_data(), guard the unbalancedness
computation against division by zero: after computing mean (from nnz and
n_constraints) check if mean is > 0 (or sufficiently > epsilon) before computing
unbalancedness = nnz_stddev / mean; if mean is zero (or nearly zero) set
unbalancedness to 0 (or a defined fallback) so nnz == 0 or zero-mean rows don't
produce NaN/inf for unbalancedness. Ensure this logic is applied where
nnz_stddev and mean are used (symbols: compute_auxiliary_data, nnz_stddev,
unbalancedness, mean, n_constraints, nnz).
- Around line 1635-1639: The static debug counters total_time_taken and
total_calls in get_problem_after_fixing_vars() are updated concurrently and
cause a data race; protect their updates by switching to a thread-safe mechanism
(e.g., std::atomic for total_calls and either std::atomic<double> or a mutex) or
add a static std::mutex and lock_guard around the increments and accumulation
before the CUOPT_LOG_TRACE call; update references to these symbols
(total_time_taken, total_calls) accordingly so their modification is
synchronized across threads.
In `@cpp/src/mip/solver.cu`:
- Around line 171-184: branch_and_bound_settings currently uses
timer_.remaining_time() which is wall‑clock based and breaks determinism when
context.settings.determinism_mode == CUOPT_MODE_DETERMINISTIC; change the logic
in the block that sets branch_and_bound_settings.time_limit so that when
determinism is requested you set a fixed deterministic limit (e.g., call
get_time_limit() or a dedicated deterministic_time_limit) or disable the time
limit (set to infinity) instead of timer_.remaining_time(); update references to
timer_.remaining_time(), branch_and_bound_settings.time_limit, and
context.settings.determinism_mode to implement this deterministic behavior.
In `@cpp/src/utilities/hashing.hpp`:
- Around line 29-44: compute_hash is marked __host__ __device__ but calls
std::memcpy (not available on device); remove the memcpy and compute the hash by
reading the bytes of val directly in a device-safe way (e.g., replace the
byte_contents + std::memcpy usage with a loop over reinterpret_cast<const
uint8_t*>(&val)[i] or otherwise directly combine byte extraction with the
hashing loop). Update compute_hash to avoid any std::memcpy or other host-only
APIs and iterate sizeof(i_t) bytes from val for the XOR/multiply steps.
In `@cpp/src/utilities/memory_instrumentation.hpp`:
- Around line 52-93: The mutable counters byte_loads and byte_stores on
memory_instrumentation_base_t are updated from const methods (reset_counters,
record_load, record_store, record_rmw) and cause data races in multithreaded
use; make them thread-safe by changing their type to std::atomic<size_t> and
update reset_counters, record_load, record_store, and record_rmw to use atomic
operations (e.g., store/fetch_add) accordingly, or alternatively document that
instrumentation is single-threaded if you choose not to change the types.
In `@cpp/src/utilities/version_info.cpp`:
- Around line 166-201: In get_cpu_max_clock_mhz() the std::stod call inside the
thread_local cached lambda can throw std::invalid_argument or std::out_of_range
during static initialization; wrap the parsing of the substring (the value after
the ':' in the "cpu MHz" line) in a try/catch that catches those exceptions (and
any std::exception as a fallback), skip malformed entries (or treat them as
zero) and continue scanning, and ensure the lambda returns 0.0 on error so
cached_mhz remains valid; update the parsing block around std::stod to trim
whitespace, check for empty substrings before calling std::stod, and catch
exceptions to avoid aborting initialization in get_cpu_max_clock_mhz().
In `@cpp/src/utilities/work_limit_timer.hpp`:
- Around line 133-140: elapsed_time() can dereference work_context when
deterministic is true; update the function to check work_context before using it
(e.g., if (deterministic && work_context) return
work_context->global_work_units_elapsed - work_units_at_start; else return
timer.elapsed_time() or a safe default) so it matches the null-checking behavior
used in check_limit() and remaining_units(); adjust to use the same fallback
behavior those methods use to avoid null-pointer access.
In `@cpp/src/utilities/work_unit_scheduler.cpp`:
- Around line 43-55: The barrier can deadlock when a context is removed while
others are waiting; change the barrier logic in work_unit_scheduler_t so waiting
threads wake when the remaining participant count matches contexts_at_barrier_.
Specifically: update the wait predicate that currently checks
"barrier_generation_ != my_generation || stopped_.load()" to also return true
when "contexts_at_barrier_ == contexts_.size()"; and in
work_unit_scheduler_t::deregister_context (and any code that mutates contexts_)
after erasing the context do a cv_.notify_all() and, if contexts_at_barrier_ ==
contexts_.size(), increment barrier_generation_ (or otherwise advance the
barrier) so the predicate becomes true for sleepers. Reference symbols:
work_unit_scheduler_t::deregister_context, contexts_, contexts_at_barrier_,
barrier_generation_, stopped_, cv_.notify_all().
In `@cpp/tests/mip/determinism_test.cu`:
- Around line 168-175: Remove the redundant explicit calls to
cuopt::seed_generator::set_seed(seed) before each solve; settings.seed is
already set (settings.seed = seed) and solve_mip(&handle_, problem, settings)
will internally call seed_generator::set_seed(settings.seed) when settings.seed
>= 0, so delete the extra cuopt::seed_generator::set_seed(seed) invocations
around the three calls to solve_mip to rely on the solver's internal seeding and
keep the test deterministic.
♻️ Duplicate comments (7)
cpp/src/mip/diversity/diversity_manager.cu (1)
337-345: GPU-heuristics-disabled path can still wait forever.This loop lacks a time-limit guard. If B&B never preempts, the solver can hang. Please add a timeout similar to the deterministic branch.
🛠️ Proposed fix (timeout guard)
- while (!check_b_b_preemption()) { - std::this_thread::sleep_for(std::chrono::milliseconds(100)); - } + while (!check_b_b_preemption()) { + if (timer.check_time_limit()) { + CUOPT_LOG_WARN("Timed out waiting for B&B preemption (GPU heuristics disabled)"); + break; + } + std::this_thread::sleep_for(std::chrono::milliseconds(100)); + }cpp/src/utilities/work_unit_predictor.hpp (1)
32-38: Guard against invalid CPU clock readings.
max_clockcan be negative or non‑finite; the current check only handles exactly zero.🐛 Proposed fix
-#include <cstdint> +#include <cstdint> +#include <cmath> ... - double max_clock = get_cpu_max_clock_mhz(); - scaling_factor_ = max_clock == 0 ? 1 : baseline_max_clock / max_clock; + const double max_clock = get_cpu_max_clock_mhz(); + scaling_factor_ = + (std::isfinite(max_clock) && max_clock > 0.0) ? (baseline_max_clock / max_clock) : 1.0;cpp/src/dual_simplex/pseudo_costs.hpp (1)
44-54: Non‑finite pseudo‑cost sums skew averages.
Counts increment even when the sum is non‑finite, which biases averages toward 0 instead of falling back to defaults.Based on learnings, validate algorithm correctness in branching decisions.🐛 Proposed fix
- for (i_t j = 0; j < n; ++j) { - if (pc_num_down[j] > 0) { - ++num_initialized_down; - if (std::isfinite(pc_sum_down[j])) { - pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j]; - } - } - if (pc_num_up[j] > 0) { - ++num_initialized_up; - if (std::isfinite(pc_sum_up[j])) { pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; } - } - } + for (i_t j = 0; j < n; ++j) { + if (pc_num_down[j] > 0 && std::isfinite(pc_sum_down[j])) { + ++num_initialized_down; + pseudo_cost_down_avg += pc_sum_down[j] / pc_num_down[j]; + } + if (pc_num_up[j] > 0 && std::isfinite(pc_sum_up[j])) { + ++num_initialized_up; + pseudo_cost_up_avg += pc_sum_up[j] / pc_num_up[j]; + } + }cpp/src/utilities/work_unit_predictor.cpp (1)
38-77: Thread safety concern for static predictor instances remains unaddressed.Per past review, the static predictor instance in
branch_and_bound.cppis shared across OpenMP threads without synchronization. While the caching code is currently commented out (lines 54-63, 73), thepredict_scalarmethod itself should be thread-safe sincemodel_t::predictappears to be stateless. However, if caching is re-enabled, synchronization will be required.cpp/src/utilities/work_unit_scheduler.cpp (1)
62-68: Synchronizebarrier_generation_in verbose logging.
Line 64 readsbarrier_generation_without the mutex whilewait_at_sync_point()mutates it under lock, so this is still a data race.🔧 Suggested fix (capture generation + target under the mutex)
- if (verbose) { - double sync_target = current_sync_target(); - CUOPT_LOG_DEBUG("[%s] Work recorded: %f, sync_target: %f (gen %zu)", - ctx.name.c_str(), - total_work, - sync_target, - barrier_generation_); - } + if (verbose) { + size_t gen = 0; + double sync_target = 0.0; + { + std::lock_guard<std::mutex> lock(mutex_); + gen = barrier_generation_; + sync_target = (sync_interval_ <= 0) + ? std::numeric_limits<double>::infinity() + : (barrier_generation_ + 1) * sync_interval_; + } + CUOPT_LOG_DEBUG("[%s] Work recorded: %f, sync_target: %f (gen %zu)", + ctx.name.c_str(), + total_work, + sync_target, + gen); + }As per coding guidelines, avoid thread-unsafe shared state.
cpp/src/dual_simplex/branch_and_bound.cpp (1)
2728-2760: Repair queue is still skipped in BSP sync.The TODO block still prevents
repair_queue_from being drained during deterministic sync, so infeasible injected solutions remain unprocessed.cpp/src/dual_simplex/bsp_debug.hpp (1)
26-26: Remove debug/attribution comment.The comment
// Thank you Cursor!appears to be a leftover from development and should be removed before merging. (This was flagged in a previous review.)
🧹 Nitpick comments (17)
cpp/src/mip/utilities/cpu_worker_thread.cuh (1)
135-143: Prefersteady_clockand avoid logging while holding the mutex.This measures elapsed time and logs inside the critical section. Consider using
std::chrono::steady_clockfor duration measurement and moving the log outside the lock to minimize lock hold time while preserving the lock forget_result()if needed.♻️ Proposed refactor
- auto wait_start = std::chrono::high_resolution_clock::now(); - std::unique_lock<std::mutex> lock(cpu_mutex); + std::unique_lock<std::mutex> lock(cpu_mutex); + auto wait_start = std::chrono::steady_clock::now(); cpu_cv.wait(lock, [this] { return cpu_thread_done || cpu_thread_terminate; }); - auto wait_end = std::chrono::high_resolution_clock::now(); - double wait_time = std::chrono::duration<double>(wait_end - wait_start).count(); - if (wait_time > 1.0) { CUOPT_LOG_DEBUG("CPU thread wait time: %.2f seconds", wait_time); } - - return static_cast<Derived*>(this)->get_result(); + auto wait_end = std::chrono::steady_clock::now(); + auto result = static_cast<Derived*>(this)->get_result(); + lock.unlock(); + double wait_time = std::chrono::duration<double>(wait_end - wait_start).count(); + if (wait_time > 1.0) { CUOPT_LOG_DEBUG("CPU thread wait time: %.2f seconds", wait_time); } + return result;cpp/src/dual_simplex/node_queue.hpp (1)
62-63: Documentdata()lifetime/thread-safety expectations.The returned const reference becomes invalid after any heap mutation and isn’t synchronized. Please document that callers must hold the queue lock and not retain the reference (or consider returning a snapshot/span). As per coding guidelines, ensure proper synchronization of shared state.
cpp/src/dual_simplex/vector_math.hpp (1)
86-95: Consider constraining VectorI_out to requireresize()support.The
inverse_permutationfunction callspinv.resize(n)on line 90, which assumesVectorI_outsupports this operation. If used with a fixed-size array or span, this will fail at compile time with an unclear error.Consider adding a static assertion or documenting the requirement:
static_assert(requires { pinv.resize(n); }, "VectorI_out must support resize()");Or alternatively, require the caller to pre-size the output and use an assertion like the other functions.
cpp/src/dual_simplex/dual_simplex_features.hpp (1)
119-148: Consider using conditional compilation for debug logging.The
log_features()method has all its printf code commented out. Since#define CUOPT_DEBUG_WORK_PREDICTIONexists at line 18, consider wrapping the logging code with that macro instead of commenting it out:void log_features(const simplex_solver_settings_t<i_t, f_t>& /* settings */) const { `#ifdef` CUOPT_DEBUG_WORK_PREDICTION printf("DS_FEATURES: iter=%d m=%d ...", iteration, ...); `#endif` }This approach maintains compile-time verification of the format string and arguments.
cpp/src/utilities/work_limit_timer.hpp (1)
174-185: Consider using CUOPT_LOG_DEBUG instead of printf for consistency.The
print_debug()method uses rawprintfwhile other debug output in this file usesCUOPT_LOG_DEBUG. Using the logging macro would ensure consistent log formatting and respect log level settings.♻️ Proposed refactor
void print_debug(std::string msg) const { if (deterministic) { - printf("%s work_limit: %f remaining_work: %f elapsed_work: %f \n", - msg.c_str(), - work_limit, - remaining_time(), - elapsed_time()); + CUOPT_LOG_DEBUG("%s work_limit: %f remaining_work: %f elapsed_work: %f", + msg.c_str(), + work_limit, + remaining_time(), + elapsed_time()); } else { timer.print_debug(msg); } }cpp/src/utilities/work_unit_scheduler.hpp (1)
35-79: Guard scheduler state from unsynchronized writes
verboseandsync_interval_are public, so callers can mutate them without the scheduler’s mutex, which can introduce data races once workers are active. Consider making them private and updating under the scheduler lock (or via atomics) through accessors.cpp/src/dual_simplex/solve.cpp (1)
193-205: Consider clarifying the NVTX scope name for Phase 1.The scope is named
"DualSimplex::phase2"because it callsdual_phase2(), but contextually this is Phase 1 of the simplex method (finding a dual-feasible solution). While technically accurate for the function name, consider"DualSimplex::phase1_dual"or similar to avoid confusion when reading traces.cpp/tests/linear_programming/c_api_tests/c_api_test.c (2)
1256-1257: Consider documenting or seedingrand()behavior.
rand()without priorsrand()seeding will produce the same sequence on each program invocation (typically starts from seed 1). If this is intentional for reproducible test runs, consider adding a comment. Otherwise, consider seeding withsrand(time(NULL))or accepting the seed as a parameter.
1310-1318: Add a clarifying comment about why exact equality is required for determinism testing.The exact equality check is appropriate for validating determinism—bitwise-identical results are the defining feature being tested. Since the code explicitly enables
CUOPT_MODE_DETERMINISTIC, this comparison is intentional and correct.A brief comment explaining that exact equality is required (not a tolerance-worthy comparison) would prevent future developers from misinterpreting the logic:
/* Determinism mode guarantees bitwise-identical results; exact equality is required */ if (first_objective != objective) { // ... }Note: Tolerance-based comparisons would be inappropriate here, as they would weaken the determinism test itself.
cpp/tests/mip/determinism_test.cu (2)
101-104: Remove stray semicolon.There's an extraneous semicolon on line 102.
♻️ Proposed fix
auto seed = std::random_device{}() & 0x7fffffff; - ; std::cout << "Tested with seed " << seed << "\n";
159-160: Remove stray semicolon.There's an extraneous semicolon on line 160.
♻️ Proposed fix
auto seed = std::random_device{}() & 0x7fffffff; - ; std::cout << "Tested with seed " << seed << "\n";cpp/src/dual_simplex/mip_node.hpp (1)
311-321: Consider epsilon comparison for floating-point lower_bound.The comparator uses direct inequality (
a.lower_bound != b.lower_bound) for floating-point comparison. While this is typical for priority queues (where exact bit equality determines ordering), iflower_boundvalues are computed via different paths, slight floating-point differences could cause non-deterministic ordering.For BSP determinism, this should be acceptable since bounds are computed identically across runs, but worth noting if issues arise.
cpp/src/utilities/work_unit_predictor.cpp (1)
44-52: Eliminate redundant map lookup.The code calls
features.find()and thenfeatures.at()with the same key, performing two lookups. Use the iterator fromfind()directly.♻️ Proposed fix
typename model_t::Entry data[model_t::NUM_FEATURES]; for (int i = 0; i < model_t::NUM_FEATURES; ++i) { - if (features.find(std::string(model_t::feature_names[i])) == features.end()) { + auto it = features.find(std::string(model_t::feature_names[i])); + if (it == features.end()) { data[i].missing = -1; CUOPT_LOG_WARN("Feature %s: missing\n", model_t::feature_names[i]); } else { - data[i].fvalue = features.at(std::string(model_t::feature_names[i])); + data[i].fvalue = it->second; } }cpp/src/mip/solve.cu (1)
144-146: Doublegetenvcall is inefficient.
std::getenv("CUOPT_MIP_HIDE_SOLUTION")is called twice. Cache the result in a single call.♻️ Proposed fix
- int hidesol = - std::getenv("CUOPT_MIP_HIDE_SOLUTION") ? atoi(std::getenv("CUOPT_MIP_HIDE_SOLUTION")) : 0; + const char* hide_env = std::getenv("CUOPT_MIP_HIDE_SOLUTION"); + int hidesol = hide_env ? atoi(hide_env) : 0; if (!hidesol) { detail::print_solution(scaled_problem.handle_ptr, sol.get_solution()); }cpp/src/dual_simplex/right_looking_lu.hpp (1)
27-33: Consider template consistency between LU variants.
right_looking_luuses theVectorItemplate forqandpinv, butright_looking_lu_row_permutation_onlyuses hardcodedstd::vector<i_t>&. All current call sites passstd::vector<i_t>, so templating this function is not strictly necessary; however, aligning the template signature for consistency would reduce API surface fragmentation if instrumentation or otherVectorItypes (e.g.,ins_vector) are introduced in the future.cpp/src/dual_simplex/bb_event.hpp (1)
143-153: Default-initialize batch horizon bounds.
horizon_start/horizon_endare uninitialized unlessclear()is called. Consider in-class defaults to avoid accidental use of indeterminate values.♻️ Suggested tweak
- double horizon_start; - double horizon_end; + double horizon_start = 0.0; + double horizon_end = 0.0;cpp/src/utilities/memory_instrumentation.hpp (1)
406-416: Direct counter access bypasses any future synchronization.Line 410 directly increments
wrapper_->byte_loadsinstead of usingrecord_load<value_type>(). If the counter is later made atomic, this code path would need updating.♻️ Suggested change for consistency
auto operator*() const { if constexpr (IsConst) { -#ifdef CUOPT_ENABLE_MEMORY_INSTRUMENTATION - wrapper_->byte_loads += sizeof(value_type); -#endif + wrapper_->template record_load<value_type>(); return *iter_; } else { return element_proxy_t(*iter_, *wrapper_); } }
| #include <algorithm> | ||
| #include <cstdint> | ||
| #include <tuple> | ||
| #include <vector> |
There was a problem hiding this comment.
Missing <limits> include breaks compilation.
std::numeric_limits is used in make_infeasible/make_numerical, but <limits> isn’t included in this header.
🔧 Suggested fix
`#include` <algorithm>
`#include` <cstdint>
+#include <limits>
`#include` <tuple>
`#include` <vector>🧰 Tools
🪛 Clang (14.0.6)
[error] 10-10: 'algorithm' file not found
(clang-diagnostic-error)
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/bb_event.hpp` around lines 10 - 13, The header uses
std::numeric_limits in make_infeasible and make_numerical but does not include
<limits>, causing compilation failures; add the missing `#include` <limits> to
cpp/src/dual_simplex/bb_event.hpp (near the other standard includes) so that
std::numeric_limits references in the make_infeasible and make_numerical
implementations resolve correctly.
| #include <optional> | ||
|
|
||
| #include <cmath> | ||
| #include <deque> | ||
| #include <limits> | ||
| #include <memory> | ||
| #include <queue> | ||
| #include <vector> |
There was a problem hiding this comment.
Avoid overflow in next_creation_seq for large B&B trees.
An int32_t counter can wrap on big instances, causing deterministic ID collisions and unstable ordering. Consider widening to 64-bit.
🛠️ Proposed fix
+#include <cstdint>
@@
- int32_t next_creation_seq{0};
+ int64_t next_creation_seq{0};Also applies to: 114-116
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/bb_worker_state.hpp` around lines 20 - 27, The counter
next_creation_seq in bb_worker_state.hpp can overflow as an int32_t; change its
type to a 64-bit integer (e.g., int64_t or uint64_t) wherever it is declared and
used (including the other related counters referenced around lines 114-116) to
ensure unique, monotonic IDs for large B&B trees; update any constructors,
accessors, comparisons, and serialization code that reference next_creation_seq
(and the other identified sequence/counter members) to use the new 64-bit type
so behavior and ordering remain stable.
| void reset_for_horizon(double horizon_start, double horizon_end, f_t global_upper_bound) | ||
| { | ||
| clock = horizon_start; | ||
| work_context.global_work_units_elapsed = horizon_start; | ||
| events.clear(); | ||
| events.horizon_start = horizon_start; | ||
| events.horizon_end = horizon_end; | ||
| event_sequence = 0; | ||
| nodes_processed_this_horizon = 0; | ||
| local_upper_bound = global_upper_bound; | ||
| local_lower_bound_ceiling = std::numeric_limits<f_t>::infinity(); | ||
| integer_solutions.clear(); | ||
| pseudo_cost_updates.clear(); | ||
| } |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
# First, let's examine the bb_worker_state.hpp file to understand the class structure
cat -n cpp/src/dual_simplex/bb_worker_state.hpp | head -200Repository: NVIDIA/cuopt
Length of output: 8765
🏁 Script executed:
# Search for horizon_start and horizon_end to see where they are used/declared
rg "horizon_start|horizon_end" cpp/src/dual_simplex/bb_worker_state.hpp -nRepository: NVIDIA/cuopt
Length of output: 1382
🏁 Script executed:
# Look for the set_snapshots method mentioned in the review
rg "set_snapshots" cpp/src/dual_simplex/bb_worker_state.hpp -n -A 5Repository: NVIDIA/cuopt
Length of output: 740
Synchronize horizon state in reset_for_horizon with member variables.
reset_for_horizon receives horizon_start and horizon_end parameters but doesn't update the member variables at lines 111-112. Currently, these values are stored only in the events batch, while the member variables horizon_start and horizon_end retain stale values until set_snapshots() is called. This creates a state inconsistency during the reset phase. Add assignments to ensure the member variables are synchronized:
void reset_for_horizon(double horizon_start, double horizon_end, f_t global_upper_bound)
{
clock = horizon_start;
+ this->horizon_start = horizon_start;
+ this->horizon_end = horizon_end;
work_context.global_work_units_elapsed = horizon_start;🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/bb_worker_state.hpp` around lines 173 - 186,
reset_for_horizon currently updates events.horizon_start and events.horizon_end
but not the worker state's member variables horizon_start and horizon_end,
leaving them stale until set_snapshots() runs; modify reset_for_horizon to
assign the incoming horizon_start and horizon_end parameters to the class
members (horizon_start and horizon_end) in addition to setting
events.horizon_start/events.horizon_end so state is synchronized immediately
(keep other resets like event_sequence, clock, and work_context changes intact).
| void initialize(int num_workers, | ||
| const std::vector<bnb_worker_type_t>& diving_types, | ||
| const lp_problem_t<i_t, f_t>& original_lp, | ||
| const csr_matrix_t<i_t, f_t>& Arow, | ||
| const std::vector<variable_type_t>& var_types, | ||
| i_t refactor_frequency, | ||
| bool deterministic) | ||
| { | ||
| workers_.clear(); | ||
| workers_.reserve(num_workers); | ||
| for (int i = 0; i < num_workers; ++i) { | ||
| bnb_worker_type_t type = diving_types[i % diving_types.size()]; | ||
| workers_.emplace_back(i, type); | ||
| workers_.back().initialize(original_lp, Arow, var_types, refactor_frequency, deterministic); | ||
| } |
There was a problem hiding this comment.
Guard against empty diving_types.
If diving_types is empty, i % diving_types.size() will crash. Provide a default type or explicit precondition.
🛠️ Proposed fix
- for (int i = 0; i < num_workers; ++i) {
- bnb_worker_type_t type = diving_types[i % diving_types.size()];
+ for (int i = 0; i < num_workers; ++i) {
+ bnb_worker_type_t type = diving_types.empty()
+ ? bnb_worker_type_t::PSEUDOCOST_DIVING
+ : diving_types[i % diving_types.size()];
workers_.emplace_back(i, type);
workers_.back().initialize(original_lp, Arow, var_types, refactor_frequency, deterministic);
}📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| void initialize(int num_workers, | |
| const std::vector<bnb_worker_type_t>& diving_types, | |
| const lp_problem_t<i_t, f_t>& original_lp, | |
| const csr_matrix_t<i_t, f_t>& Arow, | |
| const std::vector<variable_type_t>& var_types, | |
| i_t refactor_frequency, | |
| bool deterministic) | |
| { | |
| workers_.clear(); | |
| workers_.reserve(num_workers); | |
| for (int i = 0; i < num_workers; ++i) { | |
| bnb_worker_type_t type = diving_types[i % diving_types.size()]; | |
| workers_.emplace_back(i, type); | |
| workers_.back().initialize(original_lp, Arow, var_types, refactor_frequency, deterministic); | |
| } | |
| void initialize(int num_workers, | |
| const std::vector<bnb_worker_type_t>& diving_types, | |
| const lp_problem_t<i_t, f_t>& original_lp, | |
| const csr_matrix_t<i_t, f_t>& Arow, | |
| const std::vector<variable_type_t>& var_types, | |
| i_t refactor_frequency, | |
| bool deterministic) | |
| { | |
| workers_.clear(); | |
| workers_.reserve(num_workers); | |
| for (int i = 0; i < num_workers; ++i) { | |
| bnb_worker_type_t type = diving_types.empty() | |
| ? bnb_worker_type_t::PSEUDOCOST_DIVING | |
| : diving_types[i % diving_types.size()]; | |
| workers_.emplace_back(i, type); | |
| workers_.back().initialize(original_lp, Arow, var_types, refactor_frequency, deterministic); | |
| } |
🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/bb_worker_state.hpp` around lines 513 - 527, The
initialize method must guard against an empty diving_types vector to avoid
division-by-zero when computing i % diving_types.size(); add an early check in
initialize that validates !diving_types.empty() and if empty either throw a
clear std::invalid_argument (e.g., "initialize: diving_types must not be empty")
or set a documented default bnb_worker_type_t before the loop; ensure the
check/run uses the symbols initialize, diving_types, workers_, and
bnb_worker_type_t so the fix is applied right before the for-loop that currently
does i % diving_types.size().
| const int total_thread_count = num_bfs_workers + num_diving_workers; | ||
|
|
||
| #pragma omp parallel num_threads(total_thread_count) | ||
| { | ||
| int thread_id = omp_get_thread_num(); | ||
| if (thread_id < num_bfs_workers) { | ||
| auto& worker = (*bsp_workers_)[thread_id]; | ||
| f_t worker_start_time = tic(); | ||
| run_worker_loop(worker, search_tree_); | ||
| worker.total_runtime += toc(worker_start_time); | ||
| } else { | ||
| int diving_id = thread_id - num_bfs_workers; | ||
| auto& worker = (*bsp_diving_workers_)[diving_id]; | ||
| f_t worker_start_time = tic(); | ||
| run_diving_worker_loop(worker); | ||
| worker.total_runtime += toc(worker_start_time); | ||
| } |
There was a problem hiding this comment.
Prevent null deref when diving heuristics are disabled.
total_thread_count uses num_diving_workers even if bsp_diving_workers_ was never created (e.g., all diving heuristics disabled). The else branch then dereferences a null bsp_diving_workers_. Use the actual diving worker count and guard the else path.
🛠️ Suggested fix
- const int total_thread_count = num_bfs_workers + num_diving_workers;
+ const int actual_diving_workers = bsp_diving_workers_ ? (int)bsp_diving_workers_->size() : 0;
+ const int total_thread_count = num_bfs_workers + actual_diving_workers;
...
- } else {
+ } else if (bsp_diving_workers_) {
int diving_id = thread_id - num_bfs_workers;
auto& worker = (*bsp_diving_workers_)[diving_id];
f_t worker_start_time = tic();
run_diving_worker_loop(worker);
worker.total_runtime += toc(worker_start_time);
}🤖 Prompt for AI Agents
In `@cpp/src/dual_simplex/branch_and_bound.cpp` around lines 1968 - 1984,
total_thread_count currently adds num_diving_workers even when
bsp_diving_workers_ may be null, causing a null deref in the else branch;
compute the actual diving count from bsp_diving_workers_ (e.g.,
actual_diving_count = bsp_diving_workers_ ? bsp_diving_workers_->size() : 0),
use total_thread_count = num_bfs_workers + actual_diving_count, and in the
parallel region guard the diving path by checking diving_id <
actual_diving_count (or only spawn diving threads when actual_diving_count > 0)
before accessing bsp_diving_workers_ and calling run_diving_worker_loop; update
references to thread_id, num_bfs_workers, diving_id, bsp_diving_workers_,
run_diving_worker_loop and total_thread_count accordingly.
| struct memory_instrumentation_base_t { | ||
| #if CUOPT_ENABLE_MEMORY_INSTRUMENTATION | ||
| HDI void reset_counters() const { byte_loads = byte_stores = 0; } | ||
|
|
||
| template <typename T> | ||
| HDI void record_load() const | ||
| { | ||
| byte_loads += sizeof(T); | ||
| } | ||
|
|
||
| template <typename T> | ||
| HDI void record_store() const | ||
| { | ||
| byte_stores += sizeof(T); | ||
| } | ||
|
|
||
| template <typename T> | ||
| HDI void record_rmw() const | ||
| { | ||
| byte_loads += sizeof(T); | ||
| byte_stores += sizeof(T); | ||
| } | ||
|
|
||
| mutable size_t byte_loads{0}; | ||
| mutable size_t byte_stores{0}; | ||
| #else | ||
| // No-op methods when instrumentation is disabled - these inline away to zero overhead | ||
| HDI void reset_counters() const {} | ||
| template <typename T> | ||
| HDI void record_load() const | ||
| { | ||
| } | ||
| template <typename T> | ||
| HDI void record_store() const | ||
| { | ||
| } | ||
| template <typename T> | ||
| HDI void record_rmw() const | ||
| { | ||
| } | ||
| #endif // CUOPT_ENABLE_MEMORY_INSTRUMENTATION | ||
| }; |
There was a problem hiding this comment.
Thread safety concern: mutable counters without synchronization.
The byte_loads and byte_stores counters (lines 75-76) are mutable and updated from const methods without synchronization. If an instrumented container is accessed from multiple threads concurrently, this creates data races.
Consider using std::atomic<size_t> if thread-safe instrumentation is required, or document that instrumentation is single-threaded only.
🔧 Suggested fix for thread-safe counters
`#if` CUOPT_ENABLE_MEMORY_INSTRUMENTATION
- mutable size_t byte_loads{0};
- mutable size_t byte_stores{0};
+ mutable std::atomic<size_t> byte_loads{0};
+ mutable std::atomic<size_t> byte_stores{0};Note: This would require updating reset_counters(), record_load(), etc. to use atomic operations.
🤖 Prompt for AI Agents
In `@cpp/src/utilities/memory_instrumentation.hpp` around lines 52 - 93, The
mutable counters byte_loads and byte_stores on memory_instrumentation_base_t are
updated from const methods (reset_counters, record_load, record_store,
record_rmw) and cause data races in multithreaded use; make them thread-safe by
changing their type to std::atomic<size_t> and update reset_counters,
record_load, record_store, and record_rmw to use atomic operations (e.g.,
store/fetch_add) accordingly, or alternatively document that instrumentation is
single-threaded if you choose not to change the types.
| double get_cpu_max_clock_mhz() | ||
| { | ||
| // thread_local to avoid an unecessary sync inserted by the compiler | ||
| // due to the standard mandating thread-safe static local variable initialization | ||
| // the extra work here is minimal. | ||
| thread_local static double cached_mhz = []() { | ||
| // Try sysfs cpufreq interface first (returns frequency in KHz) | ||
| // FIXME: assumes all available CPUs have the same max clock as CPU0 | ||
| std::ifstream freq_file("/sys/devices/system/cpu/cpu0/cpufreq/cpuinfo_max_freq"); | ||
| if (freq_file.is_open()) { | ||
| long khz = 0; | ||
| freq_file >> khz; | ||
| if (khz > 0) { return khz / 1e3; } | ||
| } | ||
|
|
||
| // Fallback: parse /proc/cpuinfo for "cpu MHz" | ||
| std::ifstream cpuinfo("/proc/cpuinfo"); | ||
| if (!cpuinfo.is_open()) return 0.0; | ||
|
|
||
| std::string line; | ||
| double max_mhz = 0.0; | ||
| while (std::getline(cpuinfo, line)) { | ||
| if (line.find("cpu MHz") != std::string::npos) { | ||
| std::size_t colon = line.find(':'); | ||
| if (colon != std::string::npos) { | ||
| double mhz = std::stod(line.substr(colon + 1)); | ||
| if (mhz > max_mhz) { max_mhz = mhz; } | ||
| } | ||
| } | ||
| } | ||
|
|
||
| return max_mhz; | ||
| }(); | ||
|
|
||
| return cached_mhz; | ||
| } |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
# Find all usages of get_cpu_max_clock_mhz function
rg -nP --type=cpp 'get_cpu_max_clock_mhz' -B3 -A3Repository: NVIDIA/cuopt
Length of output: 1430
Guard against parsing exceptions in get_cpu_max_clock_mhz.
The std::stod call at line 191 can throw std::invalid_argument or std::out_of_range if the parsed string is malformed. While this function is only called once per thread (due to thread_local static caching), an uncaught exception during initialization could prevent the solver from starting. The return value of 0.0 is already handled safely by the caller with a ternary operator check.
🛡️ Suggested defensive fix
if (line.find("cpu MHz") != std::string::npos) {
std::size_t colon = line.find(':');
if (colon != std::string::npos) {
- double mhz = std::stod(line.substr(colon + 1));
- if (mhz > max_mhz) { max_mhz = mhz; }
+ try {
+ double mhz = std::stod(line.substr(colon + 1));
+ if (mhz > max_mhz) { max_mhz = mhz; }
+ } catch (...) {
+ // Ignore malformed lines
+ }
}
}📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| double get_cpu_max_clock_mhz() | |
| { | |
| // thread_local to avoid an unecessary sync inserted by the compiler | |
| // due to the standard mandating thread-safe static local variable initialization | |
| // the extra work here is minimal. | |
| thread_local static double cached_mhz = []() { | |
| // Try sysfs cpufreq interface first (returns frequency in KHz) | |
| // FIXME: assumes all available CPUs have the same max clock as CPU0 | |
| std::ifstream freq_file("/sys/devices/system/cpu/cpu0/cpufreq/cpuinfo_max_freq"); | |
| if (freq_file.is_open()) { | |
| long khz = 0; | |
| freq_file >> khz; | |
| if (khz > 0) { return khz / 1e3; } | |
| } | |
| // Fallback: parse /proc/cpuinfo for "cpu MHz" | |
| std::ifstream cpuinfo("/proc/cpuinfo"); | |
| if (!cpuinfo.is_open()) return 0.0; | |
| std::string line; | |
| double max_mhz = 0.0; | |
| while (std::getline(cpuinfo, line)) { | |
| if (line.find("cpu MHz") != std::string::npos) { | |
| std::size_t colon = line.find(':'); | |
| if (colon != std::string::npos) { | |
| double mhz = std::stod(line.substr(colon + 1)); | |
| if (mhz > max_mhz) { max_mhz = mhz; } | |
| } | |
| } | |
| } | |
| return max_mhz; | |
| }(); | |
| return cached_mhz; | |
| } | |
| double get_cpu_max_clock_mhz() | |
| { | |
| // thread_local to avoid an unecessary sync inserted by the compiler | |
| // due to the standard mandating thread-safe static local variable initialization | |
| // the extra work here is minimal. | |
| thread_local static double cached_mhz = []() { | |
| // Try sysfs cpufreq interface first (returns frequency in KHz) | |
| // FIXME: assumes all available CPUs have the same max clock as CPU0 | |
| std::ifstream freq_file("/sys/devices/system/cpu/cpu0/cpufreq/cpuinfo_max_freq"); | |
| if (freq_file.is_open()) { | |
| long khz = 0; | |
| freq_file >> khz; | |
| if (khz > 0) { return khz / 1e3; } | |
| } | |
| // Fallback: parse /proc/cpuinfo for "cpu MHz" | |
| std::ifstream cpuinfo("/proc/cpuinfo"); | |
| if (!cpuinfo.is_open()) return 0.0; | |
| std::string line; | |
| double max_mhz = 0.0; | |
| while (std::getline(cpuinfo, line)) { | |
| if (line.find("cpu MHz") != std::string::npos) { | |
| std::size_t colon = line.find(':'); | |
| if (colon != std::string::npos) { | |
| try { | |
| double mhz = std::stod(line.substr(colon + 1)); | |
| if (mhz > max_mhz) { max_mhz = mhz; } | |
| } catch (...) { | |
| // Ignore malformed lines | |
| } | |
| } | |
| } | |
| } | |
| return max_mhz; | |
| }(); | |
| return cached_mhz; | |
| } |
🤖 Prompt for AI Agents
In `@cpp/src/utilities/version_info.cpp` around lines 166 - 201, In
get_cpu_max_clock_mhz() the std::stod call inside the thread_local cached lambda
can throw std::invalid_argument or std::out_of_range during static
initialization; wrap the parsing of the substring (the value after the ':' in
the "cpu MHz" line) in a try/catch that catches those exceptions (and any
std::exception as a fallback), skip malformed entries (or treat them as zero)
and continue scanning, and ensure the lambda returns 0.0 on error so cached_mhz
remains valid; update the parsing block around std::stod to trim whitespace,
check for empty substrings before calling std::stod, and catch exceptions to
avoid aborting initialization in get_cpu_max_clock_mhz().
| double elapsed_time() const noexcept | ||
| { | ||
| if (deterministic) { | ||
| return work_context->global_work_units_elapsed - work_units_at_start; | ||
| } else { | ||
| return timer.elapsed_time(); | ||
| } | ||
| } |
There was a problem hiding this comment.
Potential null pointer dereference in deterministic elapsed_time().
When deterministic is true but work_context is nullptr (e.g., default-constructed then flag set), line 136 will dereference a null pointer. Other methods like check_limit() and remaining_units() guard against this, but elapsed_time() does not.
🐛 Proposed fix
double elapsed_time() const noexcept
{
if (deterministic) {
+ if (!work_context) { return 0.0; }
return work_context->global_work_units_elapsed - work_units_at_start;
} else {
return timer.elapsed_time();
}
}🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_limit_timer.hpp` around lines 133 - 140,
elapsed_time() can dereference work_context when deterministic is true; update
the function to check work_context before using it (e.g., if (deterministic &&
work_context) return work_context->global_work_units_elapsed -
work_units_at_start; else return timer.elapsed_time() or a safe default) so it
matches the null-checking behavior used in check_limit() and remaining_units();
adjust to use the same fallback behavior those methods use to avoid null-pointer
access.
| void work_unit_scheduler_t::deregister_context(work_limit_context_t& ctx) | ||
| { | ||
| std::lock_guard<std::mutex> lock(mutex_); | ||
| ctx.scheduler = nullptr; | ||
| contexts_.erase(std::remove_if(contexts_.begin(), | ||
| contexts_.end(), | ||
| [&ctx](const std::reference_wrapper<work_limit_context_t>& ref) { | ||
| return &ref.get() == &ctx; | ||
| }), | ||
| contexts_.end()); | ||
| last_sync_target_.erase(&ctx); | ||
| cv_.notify_all(); | ||
| } |
There was a problem hiding this comment.
🧩 Analysis chain
🏁 Script executed:
#!/bin/bash
# Find the work_unit_scheduler header and implementation files
fd -t f "work_unit_scheduler" --type=cpp --type=hppRepository: NVIDIA/cuopt
Length of output: 226
🏁 Script executed:
#!/bin/bash
# Examine the full work_unit_scheduler implementation
rg -n "wait_at_sync_point|barrier_generation_|stopped_" -A 5 -B 2 --type=cpp --type=hppRepository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
#!/bin/bash
# Check call sites and context around deregister_context usage
rg -n "deregister_context" -C 5 --type=cpp --type=hpp --type=cuRepository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
#!/bin/bash
# Look for the barrier predicate and synchronization mechanism
rg -n "cv_\.wait|notify_all|barrier_generation_" -A 3 -B 1 --type=cpp --type=hppRepository: NVIDIA/cuopt
Length of output: 83
🏁 Script executed:
#!/bin/bash
# Find work_unit_scheduler files
find . -name "*work_unit_scheduler*" -type fRepository: NVIDIA/cuopt
Length of output: 143
🏁 Script executed:
#!/bin/bash
# Search for the barrier and wait logic without type filters
rg -n "wait_at_sync_point" -A 10 -B 2Repository: NVIDIA/cuopt
Length of output: 4068
🏁 Script executed:
#!/bin/bash
# Search for barrier_generation and contexts_ usage
rg -n "barrier_generation_|contexts_\." -B 2 -A 2Repository: NVIDIA/cuopt
Length of output: 4721
🏁 Script executed:
#!/bin/bash
# Check deregister_context call sites
rg -n "deregister_context" -B 3 -A 3Repository: NVIDIA/cuopt
Length of output: 1618
🏁 Script executed:
#!/bin/bash
# Get the full wait_at_sync_point implementation
sed -n '110,160p' ./cpp/src/utilities/work_unit_scheduler.cppRepository: NVIDIA/cuopt
Length of output: 1757
🏁 Script executed:
#!/bin/bash
# Check if contexts_at_barrier_ is reset or decremented elsewhere
rg -n "contexts_at_barrier_" -B 2 -A 2Repository: NVIDIA/cuopt
Length of output: 2848
🏁 Script executed:
#!/bin/bash
# Verify if there are any guards or checks preventing mid-barrier deregistration
rg -n "deregister_context|on_work_recorded|wait_at_sync_point" -B 5 -A 2 | head -100Repository: NVIDIA/cuopt
Length of output: 5608
Deadlock race condition during context deregistration at barrier.
The barrier synchronization has a critical flaw. If a context is deregistered while others are blocked waiting at a barrier, the waiting threads can permanently deadlock:
- Threads A and B increment
contexts_at_barrier_and sleep on the predicatebarrier_generation_ != my_generation || stopped_.load()(line 146), waiting for all contexts to arrive. - If a third context C is deregistered before arriving,
contexts_.erase()reducescontexts_.size()while A and B are sleeping. cv_.notify_all()wakes A and B, but their predicate condition remains false (barrier_generation hasn't changed).- A and B re-sleep, now trapped:
contexts_at_barrier_ == contexts_.size()(both 2), yet no code path can wake them because the barrier increment at line 129 is never reached by sleeping threads.
The wait predicate should account for deregistration by checking contexts_at_barrier_ == contexts_.size(), or deregistration must be prevented/delayed until no context is at a barrier.
🤖 Prompt for AI Agents
In `@cpp/src/utilities/work_unit_scheduler.cpp` around lines 43 - 55, The barrier
can deadlock when a context is removed while others are waiting; change the
barrier logic in work_unit_scheduler_t so waiting threads wake when the
remaining participant count matches contexts_at_barrier_. Specifically: update
the wait predicate that currently checks "barrier_generation_ != my_generation
|| stopped_.load()" to also return true when "contexts_at_barrier_ ==
contexts_.size()"; and in work_unit_scheduler_t::deregister_context (and any
code that mutates contexts_) after erasing the context do a cv_.notify_all()
and, if contexts_at_barrier_ == contexts_.size(), increment barrier_generation_
(or otherwise advance the barrier) so the predicate becomes true for sleepers.
Reference symbols: work_unit_scheduler_t::deregister_context, contexts_,
contexts_at_barrier_, barrier_generation_, stopped_, cv_.notify_all().
|
/ok to test c5438a4 |
There was a problem hiding this comment.
Actionable comments posted: 2
🤖 Fix all issues with AI agents
In `@cpp/tests/mip/determinism_test.cu`:
- Around line 101-102: The stray standalone semicolon after the seed
initialization (the line following "auto seed = std::random_device{}() &
0x7fffffff;") should be removed; edit the determinism_test.cu snippet so only
the seed assignment remains (remove the extra ";" on the next line) to eliminate
the redundant no-op.
- Around line 159-160: There is a stray standalone semicolon after the seed
initialization (the line with "auto seed = std::random_device{}() &
0x7fffffff;"); remove the extra ";" on the following line so only the variable
initialization statement remains (keep the declaration `seed` as-is and delete
the redundant semicolon).
♻️ Duplicate comments (4)
cpp/src/mip/solver.cu (1)
171-184: Wall-clock time_limit in deterministic mode.Line 172 still uses
timer_.remaining_time()fortime_limiteven in deterministic mode. This was flagged in a previous review. Thework_limit(lines 180-184) provides deterministic termination, but if time limit triggers first, it could introduce run-to-run variance.cpp/src/dual_simplex/branch_and_bound.cpp (2)
1968-1984: Thread count still uses settings value instead of actual diving worker count.The
total_thread_countcalculation usesnum_diving_workersfrom settings even thoughbsp_diving_workers_might be null (when all diving heuristics are disabled butnum_diving_workers > 0). The correctactual_diving_workersis computed at line 1916 for logging but not used here. This can cause null pointer dereference in the else branch.
2728-2760: Repair queue processing is disabled (TODO) - dead code present.The repair queue drain logic is commented out, leaving
to_repairalways empty. The subsequent sort and iteration loop (lines 2739-2758) execute but do nothing. Either implement the repair queue support or remove the dead code block.cpp/tests/mip/determinism_test.cu (1)
170-175: Redundantseed_generator::set_seedcalls before each solve.Since
settings.seed = seedis set on line 168 andsolve_mipinternally callsseed_generator::set_seed(settings.seed)whensettings.seed >= 0, the explicit calls on lines 170, 172, and 174 are redundant. The solver will reset the seed generator to the same value on each invocation.
🧹 Nitpick comments (2)
cpp/src/mip/solver.cu (2)
26-26: Verify<cmath>is needed.
std::numeric_limits<f_t>::infinity()(line 183) comes from<limits>, not<cmath>. If<cmath>isn't used elsewhere in this file, consider removing it or replacing with the correct header if needed.
246-246: Remove commented-out debug code.The commented-out verbose flag is a debug artifact. Consider removing it or controlling verbosity through a proper debug setting.
♻️ Suggested cleanup
context.work_unit_scheduler_.register_context(branch_and_bound->get_work_unit_context()); - // context.work_unit_scheduler_.verbose = true;
| auto seed = std::random_device{}() & 0x7fffffff; | ||
| ; |
There was a problem hiding this comment.
Remove stray semicolon.
Line 102 contains a standalone semicolon that appears to be a copy-paste artifact.
Proposed fix
auto seed = std::random_device{}() & 0x7fffffff;
- ;
std::cout << "Tested with seed " << seed << "\n";📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| auto seed = std::random_device{}() & 0x7fffffff; | |
| ; | |
| auto seed = std::random_device{}() & 0x7fffffff; | |
| std::cout << "Tested with seed " << seed << "\n"; |
🤖 Prompt for AI Agents
In `@cpp/tests/mip/determinism_test.cu` around lines 101 - 102, The stray
standalone semicolon after the seed initialization (the line following "auto
seed = std::random_device{}() & 0x7fffffff;") should be removed; edit the
determinism_test.cu snippet so only the seed assignment remains (remove the
extra ";" on the next line) to eliminate the redundant no-op.
| auto seed = std::random_device{}() & 0x7fffffff; | ||
| ; |
There was a problem hiding this comment.
Remove stray semicolon.
Line 160 contains a standalone semicolon that appears to be a copy-paste artifact.
Proposed fix
auto seed = std::random_device{}() & 0x7fffffff;
- ;
std::cout << "Tested with seed " << seed << "\n";📝 Committable suggestion
‼️ IMPORTANT
Carefully review the code before committing. Ensure that it accurately replaces the highlighted code, contains no missing lines, and has no issues with indentation. Thoroughly test & benchmark the code to ensure it meets the requirements.
| auto seed = std::random_device{}() & 0x7fffffff; | |
| ; | |
| auto seed = std::random_device{}() & 0x7fffffff; | |
| std::cout << "Tested with seed " << seed << "\n"; |
🤖 Prompt for AI Agents
In `@cpp/tests/mip/determinism_test.cu` around lines 159 - 160, There is a stray
standalone semicolon after the seed initialization (the line with "auto seed =
std::random_device{}() & 0x7fffffff;"); remove the extra ";" on the following
line so only the variable initialization statement remains (keep the declaration
`seed` as-is and delete the redundant semicolon).
2 participants
This PR introduces a deterministic execution mode for the parallel branch-and-bound MIP solver.
When
determinism_mode = CUOPT_MODE_DETERMINISTIC, the solver guarantees bitwise-identical results across runs regardless of thread scheduling variations, so long as the runs are on the same platform with the same environment (= same glibc, etc...)The approach that was chosen is referred to as Bulk Synchronous Parallel (BSP) in the literature. This is an execution model where computation proceeds in discrete horizons of virtual time (=work units) separated by synchronization barriers.
To cope with the inherent nondeterminism of wall-clock-time due to factors such as caching state, OS scheduling, CPU throttling...
instead of time, progress is measured in "work units" approximating the execution time of an algorithm using known and deterministic factors, such as memory operations, problem features and sparsity...
Workers explore the tree independently and locally within a horizon, and collect events (branching decisions, solutions, pseudo-cost updates) during the process without exchanging them yet.
At each horizon boundary, events are sorted by work-unit timestamp with tie-breaking, history is replayed in deterministic order to update global state, and if need be nodes are redistributed across workers to balance load (which replaces the ramp-up mechanism).
Workers operate on local snapshots of shared state (pseudo-costs, upper bound, LP iteration count) taken at horizon start to avoid read races. This trades some accuracy in decisions (pruning may rely on a higher lower bound than in nondeterminsitic mode), for determinism.
Support for start CPU heuristics is present, based on a "producer" model: CPUFJ starts immediately and begins producing solutions with work unit timestamps, that are produced by the B&B thread once the appropriate time is reached. A form of one-way synchronization is implemented to prevent the CPUFJ thread from "falling behind" and producing solutiosn in the past from the perspective of the B&B thread. CPUFJ is allowed to run ahead, and is biased to run faster than B&B to avoid unnecessary syncs.
B&B BFS workers, diving workers, and a start CPUFJ thread are supported in deterministic mode. GPU Heuristics are disabled in this mode and will be incorporated into the deterministic framework in the future.
Note: files in the
utilities/models/folders are auto-generated from the LightGBM models trained to predict work units from algorithm features.The implementation is described in more detail in this document: B&B Determinism
Benchmark results
B&B alone, 10min
Regression testing again main branch:
Summary by CodeRabbit
New Features
Improvements
Tests
✏️ Tip: You can customize this high-level summary in your review settings.