Add NPS=4 model (#417)

* Add NPS=4 model * Fixing bug with ExtractSubDataset function not fully initializing subdataset (#390) (cherry picked from commit f6f5e16) Co-authored-by: Stanley Tsang <[email protected]>
ROCm · Aug 6, 2021 · d306fd3 · d306fd3
1 parent 9e2f016
commit d306fd3
Show file tree

Hide file tree

Showing 12 changed files with 40 additions and 13 deletions.
diff --git a/src/graph/rome_models.cc b/src/graph/rome_models.cc
@@ -286,6 +286,17 @@ static struct rcclRomeModel rome_model_43 = {
   .ringBase = "0 1 2 3 4 5 6 7|0 2 5 7 4 6 1 3|0 3 1 6 4 7 5 2|0 7 6 5 4 3 2 1",
 };
 
+static struct rcclRomeModel rome_model_53 = {
+  .nGpus = 8, .nCpus = 4, .nNics = 4, .nLinks = 3,
+  .gpuIds = { 0x4a000, 0x50000, 0xa000, 0xf000, 0xca000, 0xd0000, 0x8a000, 0x90000, },
+  .nicIds = { 0x0, 0x0, 0x0, 0x0, },
+  .gpuNuma = { 1, 1, 3, 3, 5, 5, 7, 7, },
+  .nicNuma = { 1, 3, 5, 7, },
+  .connMatrix = { 0, 1, 1, 1, 0, 0, 0, 0, 1, 0, 1, 1, 0, 0, 0, 0, 1, 1, 0, 1, 0, 0, 0, 0, 1, 1, 1, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 1, 1, 1, 0, 0, 0, 0, 1, 0, 1, 1, 0, 0, 0, 0, 1, 1, 0, 1, 0, 0, 0, 0, 1, 1, 1, 0, },
+  .pattern = "00210021",
+  .ringBase = "N0 0 1 2 3 4 5 6 7 N3|N3 7 6 5 4 3 2 1 0 N0|N1 2 3 0 1 6 7 4 5 N2|N2 5 4 7 6 1 0 3 2 N1",
+};
+
 static struct rcclRomeModel romeTopoModels[] = {
   rome_model_22,
   rome_model_25,
@@ -309,6 +320,7 @@ static struct rcclRomeModel romeTopoModels[] = {
   rome_model_49,
   rome_model_52,
   rome_model_43,
+  rome_model_53,
 };
 
 /* Parse user defined rings. Format is like :

diff --git a/test/CorrectnessTest.hpp b/test/CorrectnessTest.hpp
@@ -212,6 +212,7 @@ namespace CorrectnessTests
             subDataset.numElements = lastElement - startElement + 1;
             subDataset.dataType    = dataType;
             subDataset.inPlace     = inPlace;
+            subDataset.function    = function;
 
             subDataset.inputs.resize(numDevices);
             subDataset.outputs.resize(numDevices);

diff --git a/test/test_AllGather.cpp b/test/test_AllGather.cpp
@@ -14,7 +14,7 @@ namespace CorrectnessTests
 
         // Prepare input / output / expected results
         Dataset dataset;
-        dataset.Initialize(numDevices, numElements, dataType, inPlace);
+        dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollAllGather);
         FillDatasetWithPattern(dataset);
         ComputeExpectedResults(dataset);
 
@@ -46,7 +46,7 @@ namespace CorrectnessTests
 
         // Allocate dataset
         Dataset dataset;
-        dataset.Initialize(numDevices, numElements, dataType, inPlace);
+        dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollAllGather);
 
         // Loop over several offsets (so that device pointers are not aligned)
         for (int firstElement = 1; firstElement <= 11; firstElement += 2)

diff --git a/test/test_AllReduce.cpp b/test/test_AllReduce.cpp
@@ -14,7 +14,7 @@ namespace CorrectnessTests
 
     // Prepare input / output / expected results
     Dataset dataset;
-    dataset.Initialize(numDevices, numElements, dataType, inPlace);
+    dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollAllReduce);
     FillDatasetWithPattern(dataset);
     ComputeExpectedResults(dataset, op);
 

diff --git a/test/test_AllReduceAbort.cpp b/test/test_AllReduceAbort.cpp
@@ -30,7 +30,7 @@ namespace CorrectnessTests
 
         // Prepare input / output / expected results
         Dataset dataset;
-        dataset.Initialize(numDevices, numElements, dataType, inPlace);
+        dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollAllReduce);
         FillDatasetWithPattern(dataset);
 
         int gpu = 0; // GPU number to trigger abort

diff --git a/test/test_AllReduceGroup.cpp b/test/test_AllReduceGroup.cpp
@@ -15,9 +15,9 @@ namespace CorrectnessTests
 
     // Prepare input / output / expected results
     Dataset dataset1, dataset2, dataset3;
-    dataset1.Initialize(numDevices, numElements, dataType, inPlace);
-    dataset2.Initialize(numDevices, numElements, dataType, inPlace);
-    dataset3.Initialize(numDevices, numElements, dataType, inPlace);
+    dataset1.Initialize(numDevices, numElements, dataType, inPlace, ncclCollAllReduce);
+    dataset2.Initialize(numDevices, numElements, dataType, inPlace, ncclCollAllReduce);
+    dataset3.Initialize(numDevices, numElements, dataType, inPlace, ncclCollAllReduce);
     FillDatasetWithPattern(dataset1);
     FillDatasetWithPattern(dataset2);
     FillDatasetWithPattern(dataset3);

diff --git a/test/test_Broadcast.cpp b/test/test_Broadcast.cpp
@@ -14,7 +14,7 @@ namespace CorrectnessTests
 
         // Allocate data
         Dataset dataset;
-        dataset.Initialize(numDevices, numElements, dataType, inPlace);
+        dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollBroadcast);
 
         // Test each possible root
         for (int root = 0; root < numDevices; root++)

diff --git a/test/test_BroadcastAbort.cpp b/test/test_BroadcastAbort.cpp
@@ -30,7 +30,7 @@ namespace CorrectnessTests
 
         // Prepare input / output / expected results
         Dataset dataset;
-        dataset.Initialize(numDevices, numElements, dataType, inPlace);
+        dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollBroadcast);
         FillDatasetWithPattern(dataset);
 
         int root = 0;

diff --git a/test/test_CombinedCalls.cpp b/test/test_CombinedCalls.cpp
@@ -20,9 +20,16 @@ namespace CorrectnessTests
 
         // Create multiple datasets for combined operation
         std::vector<Dataset> datasets(5);
+        std::vector<ncclFunc_t> ncclFuncs(5);
+        ncclFuncs.push_back(ncclCollAllGather);
+        ncclFuncs.push_back(ncclCollAllReduce);
+        ncclFuncs.push_back(ncclCollBroadcast);
+        ncclFuncs.push_back(ncclCollReduce);
+        ncclFuncs.push_back(ncclCollReduceScatter);
+
         for (int i = 0; i < datasets.size(); i++)
         {
-            datasets[i].Initialize(numDevices, numElements, dataType, inPlace);
+            datasets[i].Initialize(numDevices, numElements, dataType, inPlace, ncclFuncs[i]);
             FillDatasetWithPattern(datasets[i]);
         }
 

diff --git a/test/test_GroupCalls.cpp b/test/test_GroupCalls.cpp
@@ -19,9 +19,16 @@ namespace CorrectnessTests
 
         // Create multiple datasets for group operation
         std::vector<Dataset> datasets(5);
+        std::vector<ncclFunc_t> ncclFuncs(5);
+        ncclFuncs.push_back(ncclCollAllGather);
+        ncclFuncs.push_back(ncclCollAllReduce);
+        ncclFuncs.push_back(ncclCollBroadcast);
+        ncclFuncs.push_back(ncclCollReduce);
+        ncclFuncs.push_back(ncclCollReduceScatter);
+
         for (int i = 0; i < datasets.size(); i++)
         {
-            datasets[i].Initialize(numDevices, numElements, dataType, inPlace);
+            datasets[i].Initialize(numDevices, numElements, dataType, inPlace, ncclFuncs[i]);
             FillDatasetWithPattern(datasets[i]);
         }
 

diff --git a/test/test_Reduce.cpp b/test/test_Reduce.cpp
@@ -14,7 +14,7 @@ namespace CorrectnessTests
 
         // Allocate data
         Dataset dataset;
-        dataset.Initialize(numDevices, numElements, dataType, inPlace);
+        dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollReduce);
 
         // Test each possible root
         for (int root = 0; root < numDevices; root++)

diff --git a/test/test_ReduceScatter.cpp b/test/test_ReduceScatter.cpp
@@ -15,7 +15,7 @@ namespace CorrectnessTests
 
         // Prepare input / output / expected results
         Dataset dataset;
-        dataset.Initialize(numDevices, numElements, dataType, inPlace);
+        dataset.Initialize(numDevices, numElements, dataType, inPlace, ncclCollReduceScatter);
         FillDatasetWithPattern(dataset);
         ComputeExpectedResults(dataset, op);