grt: multicore congestion handling

src/OpenRoad.cc

@@ -668,6 +668,9 @@ void OpenRoad::setThreadCount(int threads, bool print_info)
 
   // place limits on tools with threads
   sta_->setThreadCount(threads_);
+  if (global_router_ != nullptr) {
+    global_router_->setNumThreads(threads_);
+  }
 }
 
 void OpenRoad::setThreadCount(const char* threads, bool print_info)

src/grt/README.md

@@ -26,6 +26,7 @@ global_route
     [-critical_nets_percentage percent]
     [-skip_large_fanout_nets fanout]
     [-allow_congestion]
+    [-snapshot_batched_width width]
     [-verbose]
     [-start_incremental]
     [-end_incremental]
@@ -52,6 +53,7 @@ global_route
 | `-use_cugr` | This flag run GRT using CUGR as the router solver. NOTE: this is not ready for production. |
 | `-resistance_aware` | This flag enables resistance-aware layer assignment and 3D routing. NOTE: this is not ready for production. |
 | `-infinite_cap` | Enables "infinite" gcell capacity for testing purpose. NOTE: this is not recommended for production flows. |
+| `-snapshot_batched_width` | Set the semantic width of snapshot-batched routing (max batches per wave). The default is `0`, preserving the non-batched behavior. Set a positive integer to enable batched routing; allowed values are integers `[1, MAX_INT]`. Execution width still follows `set_thread_count`. NOTE: this is not recommended for production flows; it is intended for exploration and research projects. |
 
 ### Set Routing Layers
 

src/grt/include/grt/GlobalRouter.h

@@ -155,12 +155,16 @@ class GlobalRouter
   void setGridOrigin(int x, int y);
   void setAllowCongestion(bool allow_congestion);
   void setResistanceAware(bool resistance_aware);
+  void setSnapshotBatchedWidth(int snapshot_batched_width);
+  int getSnapshotBatchedWidth() const;
+  int getSnapshotBatchCount() const;
   void setMacroExtension(int macro_extension);
   void setUseCUGR(bool use_cugr) { use_cugr_ = use_cugr; };
   void setSkipLargeFanoutNets(int skip_large_fanout)
   {
     skip_large_fanout_ = skip_large_fanout;
   };
+  void setNumThreads(int num_threads);
 
   void setInfiniteCapacity(bool infinite_capacity);
 
@@ -531,6 +535,8 @@ class GlobalRouter
   int congestion_report_iter_step_;
   bool allow_congestion_;
   bool resistance_aware_{false};
+  int snapshot_batched_width_{0};
+  int num_threads_;
   std::vector<int> vertical_capacities_;
   std::vector<int> horizontal_capacities_;
   int macro_extension_;

src/grt/src/GlobalRouter.cpp

@@ -87,6 +87,7 @@ GlobalRouter::GlobalRouter(utl::Logger* logger,
       adjustment_(0.0),
       congestion_report_iter_step_(0),
       allow_congestion_(false),
+      num_threads_(1),
       macro_extension_(0),
       initialized_(false),
       total_diodes_count_(0),
@@ -110,6 +111,28 @@ GlobalRouter::GlobalRouter(utl::Logger* logger,
   cugr_ = new CUGR(db_, logger_, service_registry_, stt_builder_, sta_);
 }
 
+void GlobalRouter::setNumThreads(int num_threads)
+{
+  num_threads_ = num_threads;
+  fastroute_->setNumThreads(num_threads_);
+}
+
+void GlobalRouter::setSnapshotBatchedWidth(int snapshot_batched_width)
+{
+  snapshot_batched_width_ = snapshot_batched_width;
+  fastroute_->setSnapshotBatchedWidth(snapshot_batched_width_);
+}
+
+int GlobalRouter::getSnapshotBatchedWidth() const
+{
+  return fastroute_->getSnapshotBatchedWidth();
+}
+
+int GlobalRouter::getSnapshotBatchCount() const
+{
+  return fastroute_->getSnapshotBatchCount();
+}
+
 void GlobalRouter::initGui(
     gui::HeatMapSourceHandle routing_congestion_data_source,
     gui::HeatMapSourceHandle routing_congestion_data_source_rudy)
@@ -2317,6 +2340,7 @@ void GlobalRouter::configFastRoute()
   fastroute_->setOverflowIterations(congestion_iterations_);
   fastroute_->setCongestionReportIterStep(congestion_report_iter_step_);
   fastroute_->setResistanceAware(resistance_aware_);
+  fastroute_->setSnapshotBatchedWidth(snapshot_batched_width_);
 
   if (congestion_file_name_ != nullptr) {
     fastroute_->setCongestionReportFile(congestion_file_name_);
@@ -3838,9 +3862,17 @@ int GlobalRouter::computeNetWirelength(odb::dbNet* db_net)
 
 void GlobalRouter::computeWirelength()
 {
+  std::vector<odb::dbNet*> routed_nets;
+  routed_nets.reserve(routes_.size());
+  for (const auto& [db_net, route] : routes_) {
+    routed_nets.push_back(db_net);
+  }
+
   int64_t total_wirelength = 0;
-  for (auto& net_route : routes_) {
-    total_wirelength += computeNetWirelength(net_route.first);
+#pragma omp parallel for num_threads(num_threads_) \
+    reduction(+ : total_wirelength)
+  for (int i = 0; i < static_cast<int>(routed_nets.size()); i++) {
+    total_wirelength += computeNetWirelength(routed_nets[i]);
   }
   logger_->metric("global_route__wirelength",
                   total_wirelength / block_->getDefUnits());

src/grt/src/GlobalRouter.i

@@ -110,6 +110,24 @@ set_resistance_aware(bool resistance_aware)
   getGlobalRouter()->setResistanceAware(resistance_aware);
 }
 
+void
+set_snapshot_batched_width(int snapshot_batched_width)
+{
+  getGlobalRouter()->setSnapshotBatchedWidth(snapshot_batched_width);
+}
+
+int
+get_snapshot_batched_width()
+{
+  return getGlobalRouter()->getSnapshotBatchedWidth();
+}
+
+int
+get_snapshot_batch_count()
+{
+  return getGlobalRouter()->getSnapshotBatchCount();
+}
+
 void
 set_critical_nets_percentage(float criticalNetsPercentage)
 {
@@ -178,6 +196,8 @@ void end_incremental()
 void
 global_route()
 {
+  const int num_threads = ord::OpenRoad::openRoad()->getThreadCount();
+  getGlobalRouter()->setNumThreads(num_threads);
   getGlobalRouter()->globalRoute(true);
 }
 

src/grt/src/GlobalRouter.tcl

@@ -148,6 +148,7 @@ sta::define_cmd_args "global_route" {[-guide_file out_file] \
                                   [-critical_nets_percentage percent] \
                                   [-skip_large_fanout_nets fanout] \
                                   [-allow_congestion] \
+                                  [-snapshot_batched_width width] \
                                   [-verbose] \
                                   [-start_incremental] \
                                   [-end_incremental] \
@@ -159,8 +160,8 @@ sta::define_cmd_args "global_route" {[-guide_file out_file] \
 proc global_route { args } {
   sta::parse_key_args "global_route" args \
     keys {-guide_file -congestion_iterations -congestion_report_file \
-          -grid_origin -critical_nets_percentage -congestion_report_iter_step\
-          -skip_large_fanout_nets
+          -grid_origin -critical_nets_percentage -congestion_report_iter_step \
+          -skip_large_fanout_nets -snapshot_batched_width
          } \
     flags {-allow_congestion -resistance_aware -infinite_cap -verbose -start_incremental \
           -end_incremental -use_cugr}
@@ -229,6 +230,15 @@ proc global_route { args } {
   set resistance_aware [info exists flags(-resistance_aware)]
   grt::set_resistance_aware $resistance_aware
 
+  if { [info exists keys(-snapshot_batched_width)] } {
+    set snapshot_batched_width $keys(-snapshot_batched_width)
+    sta::check_positive_integer "-snapshot_batched_width" \
+      $snapshot_batched_width
+  } else {
+    set snapshot_batched_width 0
+  }
+  grt::set_snapshot_batched_width $snapshot_batched_width
+
   set infinite_cap [info exists flags(-infinite_cap)]
   grt::set_infinite_cap $infinite_cap
 

src/grt/src/fastroute/include/FastRoute.h

@@ -169,6 +169,7 @@ class FastRouteCore
   NetRouteMap run();
   int totalOverflow() const { return total_overflow_; }
   bool has2Doverflow() const { return has_2D_overflow_; }
+  int getSnapshotBatchCount() const { return snapshot_batch_count_; }
   void getBlockage(odb::dbTechLayer* layer,
                    int x,
                    int y,
@@ -244,6 +245,12 @@ class FastRouteCore
   void setCongestionReportFile(const char* congestion_file_name);
   void setGridMax(int x_max, int y_max);
   void setDetourPenalty(int penalty);
+  void setNumThreads(int num_threads) { num_threads_ = num_threads; }
+  void setSnapshotBatchedWidth(int snapshot_batched_width)
+  {
+    snapshot_batched_width_ = snapshot_batched_width;
+  }
+  int getSnapshotBatchedWidth() const { return snapshot_batched_width_; }
   void getCongestionNets(std::set<odb::dbNet*>& congestion_nets);
   void computeCongestionInformation();
   std::vector<int> getOriginalResources();
@@ -326,6 +333,24 @@ class FastRouteCore
                      int L,
                      const CostParams& cost_params,
                      float& slack_th);
+  void mazeRouteMSMDSequential(int iter,
+                               int expand,
+                               int ripup_threshold,
+                               int maze_edge_threshold,
+                               bool ordering,
+                               int via,
+                               int L,
+                               const CostParams& cost_params,
+                               float& slack_th);
+  bool runSnapshotBatchedMazeRoute(int iter,
+                                   int expand,
+                                   int ripup_threshold,
+                                   int maze_edge_threshold,
+                                   bool ordering,
+                                   int via,
+                                   int L,
+                                   const CostParams& cost_params,
+                                   float& slack_th);
   void convertToMazeroute();
   int getOverflow2D(int* maxOverflow);
   int getOverflow2Dmaze(int* maxOverflow, int* tUsage);
@@ -621,6 +646,47 @@ class FastRouteCore
   void copyBR();
   void copyRS();
   void freeRR();
+  std::vector<int> getMazeRouteNetOrder(bool ordering, float& slack_th);
+  bool hasNonSoftNdrNets() const;
+  int resolveSnapshotExecutionThreads(int work_items) const;
+  int resolveSnapshotWaveSize(int available_batch_count) const;
+  int resolveSnapshotBaseBatchSize(int net_count) const;
+  bool useSnapshotBatchRouting(int net_count) const;
+  int resolveSnapshotBatchIterationLimit(int net_count) const;
+  bool useSnapshotBatchRoutingForIteration(int iter, int net_count) const;
+  int resolveSnapshotNetsForBatch(int iter, int net_count) const;
+  std::unique_ptr<FastRouteCore> buildSnapshotBatchWorker() const;
+  void syncSnapshotBatchWorker(const FastRouteCore& snapshot,
+                               const std::vector<int>& batch_net_ids);
+  void applySnapshotBatchRoute(int net_id, StTree&& sttree);
+  void updatePlanarNetUsage(const StTree& sttree, FrNet* net, int edge_cost);
+  void resetSnapshotBatchStats();
+
+  // Timing data collected during run(), reported via reportRunMetrics().
+  struct RunTimings
+  {
+    double total = 0.0;
+    double initial_rsmt = 0.0;
+    double route_l = 0.0;
+    double congestion_rsmt = 0.0;
+    double new_route_l = 0.0;
+    double spiral = 0.0;
+    double route_z = 0.0;
+    double monotonic = 0.0;
+    double overflow_iterations = 0.0;
+    double finalization = 0.0;
+  };
+  void reportRunMetrics(const RunTimings& timings,
+                        int num_vias,
+                        int final_length);
+
+  // Returns true if the overflow loop should break due to snapshot
+  // convergence patience being exhausted.
+  bool checkSnapshotConvergence(int past_cong,
+                                int& bmfl,
+                                int& bwcnt,
+                                int iter,
+                                int snapshot_batch_count_before);
   int edgeShift(stt::Tree& t, int net);
   int edgeShiftNew(stt::Tree& t, int net);
 
@@ -635,6 +701,23 @@ class FastRouteCore
   static const int BIG_INT = 1e9;  // big integer used as infinity
   static const int HCOST = 5000;
 
+  // Snapshot-batched routing constants.
+  // Nets are partitioned into batches and routed in parallel waves against
+  // a frozen graph snapshot.  The user-supplied snapshot_batched_width_
+  // controls the maximum batches per wave (independent of thread count) and
+  // gates the minimum routable net count (2 * width) needed to attempt
+  // batching.
+  //
+  // kSnapshotLowOverflowForSerialCleanup /
+  // kSnapshotLowMaxOverflowForSerialCleanup:
+  //   overflow thresholds below which batching is disabled for the run,
+  //   because the design is already near-converged.
+  // kSnapshotCleanupPatience: how many non-improving iterations to tolerate
+  //   in the snapshot cleanup phase before breaking.
+  static constexpr int kSnapshotLowOverflowForSerialCleanup = 1500;
+  static constexpr int kSnapshotLowMaxOverflowForSerialCleanup = 32;
+  static constexpr int kSnapshotCleanupPatience = 12;
+
   int max_degree_;
   std::vector<int> cap_per_layer_;
   std::vector<int> usage_per_layer_;
@@ -647,6 +730,12 @@ class FastRouteCore
   std::string congestion_file_name_;
   std::vector<odb::dbTechLayerDir> layer_directions_;
   std::vector<odb::dbTechLayer*> db_layers_;
+  int num_threads_;
+  // When false, nets_ contains borrowed pointers from a parent
+  // FastRouteCore (snapshot-batch workers).  Workers must not outlive
+  // the parent's nets_ lifetime.
+  bool owns_nets_;
+  int snapshot_batched_width_;
   int x_range_;
   int y_range_;
 
@@ -746,6 +835,15 @@ class FastRouteCore
 
   std::vector<int> net_ids_;
 
+  // Maze 2D variables
+  std::vector<bool> pop_heap2_2D_;
+  std::vector<double*> src_heap_2D_;
+  std::vector<double*> dest_heap_2D_;
+  multi_array<double, 2> d1_2D_;
+  multi_array<double, 2> d2_2D_;
+  std::vector<bool> visited_2D_;
+  std::vector<int> queue_2D_;
+
   // Maze 3D variables
   multi_array<Direction, 3> directions_3D_;
   multi_array<int, 3> corr_edge_3D_;
@@ -756,6 +854,15 @@ class FastRouteCore
   multi_array<int, 3> d1_3D_;
   multi_array<int, 3> d2_3D_;
   multi_array<int, 3> path_len_3D_;
+  double snapshot_batch_sync_time_ = 0.0;
+  double snapshot_batch_route_time_ = 0.0;
+  double snapshot_batch_apply_time_ = 0.0;
+  int snapshot_batch_count_ = 0;
+  int snapshot_batch_net_count_ = 0;
+  int snapshot_batch_wave_count_ = 0;
+  bool snapshot_batch_disabled_for_run_ = false;
+  bool snapshot_cleanup_active_ = false;
+  bool has_non_soft_ndr_nets_ = false;
   int detour_penalty_;
 };
 

src/grt/src/fastroute/include/Graph2D.h

@@ -55,6 +55,7 @@ class Graph2D
   void init(int x_grid, int y_grid, int num_layers, utl::Logger* logger);
   void InitEstUsage();
   void InitLastUsage(int upType);
+  void copyRoutingStateFrom(const Graph2D& other, bool include_ndr_state);
   void clear();
   void clearUsed();
   bool hasEdges() const;