From 1f0f14eba2e6253423b1a58ca38989261308df6c Mon Sep 17 00:00:00 2001 From: Alex Fender Date: Fri, 26 Mar 2021 09:31:29 -0500 Subject: [PATCH] MS BFS python APIs + EgoNet updates (#1469) There are various things in this PR. Multi-Seed (MS) BFS: - API tentative - Saving research on memory feasability helper function (not in production) - Saving research on running the current BFS concurrently with streams and threads for analysis perf comparison (not in production) EgoNet: - Multithreading in EgoNet which deserializes execution and comes with mild performance improvements on large sizes - Some cleanup Authors: - Alex Fender (@afender) Approvers: - Chuck Hastings (@ChuckHastings) - @Iroy30 - Brad Rees (@BradReesWork) URL: https://github.com/rapidsai/cugraph/pull/1469 --- cpp/src/community/egonet.cu | 19 +- cpp/tests/CMakeLists.txt | 16 ++ cpp/tests/community/egonet_test.cu | 283 +++++++++++------------ cpp/tests/experimental/ms_bfs_test.cpp | 301 +++++++++++++++++++++++++ cpp/tests/experimental/streams.cu | 44 ++++ python/cugraph/__init__.py | 4 +- python/cugraph/tests/test_egonet.py | 32 +-- python/cugraph/traversal/__init__.py | 4 +- python/cugraph/traversal/ms_bfs.py | 282 +++++++++++++++++++++++ python/cugraph/utilities/utils.py | 106 +++++---- 10 files changed, 868 insertions(+), 223 deletions(-) create mode 100644 cpp/tests/experimental/ms_bfs_test.cpp create mode 100644 cpp/tests/experimental/streams.cu create mode 100644 python/cugraph/traversal/ms_bfs.py diff --git a/cpp/src/community/egonet.cu b/cpp/src/community/egonet.cu index 067d27f9a92..336a5c939b8 100644 --- a/cpp/src/community/egonet.cu +++ b/cpp/src/community/egonet.cu @@ -79,7 +79,12 @@ extract( // Streams will allocate concurrently later std::vector> reached{}; - reached.reserve(handle.get_num_internal_streams()); + reached.reserve(n_subgraphs); + for (vertex_t i = 0; i < n_subgraphs; i++) { + // Allocations and operations are attached to the worker stream + rmm::device_uvector local_reach(v, handle.get_internal_stream_view(i)); + reached.push_back(std::move(local_reach)); + } // h_source_vertex[i] is used by other streams in the for loop user_stream_view.synchronize(); @@ -87,15 +92,13 @@ extract( HighResTimer hr_timer; hr_timer.start("ego_neighbors"); #endif + +#pragma omp parallel for for (vertex_t i = 0; i < n_subgraphs; i++) { // get light handle from worker pool raft::handle_t light_handle(handle, i); auto worker_stream_view = light_handle.get_stream_view(); - // Allocations and operations are attached to the worker stream - rmm::device_uvector local_reach(v, worker_stream_view); - reached.push_back(std::move(local_reach)); - // BFS with cutoff // consider adding a device API to BFS (ie. accept source on the device) rmm::device_uvector predecessors(v, worker_stream_view); // not used @@ -149,10 +152,10 @@ extract( neighbors.resize(h_neighbors_offsets[n_subgraphs]); user_stream_view.synchronize(); - // Construct the neighboors list concurrently +// Construct the neighboors list concurrently +#pragma omp parallel for for (vertex_t i = 0; i < n_subgraphs; i++) { - raft::handle_t light_handle(handle, i); - auto worker_stream_view = light_handle.get_stream_view(); + auto worker_stream_view = handle.get_internal_stream_view(i); thrust::copy(rmm::exec_policy(worker_stream_view), reached[i].begin(), reached[i].end(), diff --git a/cpp/tests/CMakeLists.txt b/cpp/tests/CMakeLists.txt index 5571cf5f124..1dc4a5d3eaa 100644 --- a/cpp/tests/CMakeLists.txt +++ b/cpp/tests/CMakeLists.txt @@ -318,6 +318,13 @@ set(MST_TEST_SRC ConfigureTest(MST_TEST "${MST_TEST_SRC}") +################################################################################################### +# - Experimental stream tests ----------------------------------------------------- + +set(EXPERIMENTAL_STREAM_SRCS + "${CMAKE_CURRENT_SOURCE_DIR}/experimental/streams.cu") + +ConfigureTest(EXPERIMENTAL_STREAM "${EXPERIMENTAL_STREAM_SRCS}" "") ################################################################################################### # - Experimental R-mat graph generation tests ----------------------------------------------------- @@ -375,6 +382,15 @@ set(EXPERIMENTAL_BFS_TEST_SRCS ConfigureTest(EXPERIMENTAL_BFS_TEST "${EXPERIMENTAL_BFS_TEST_SRCS}") +################################################################################################### +# - Experimental BFS tests ------------------------------------------------------------------------ + +set(EXPERIMENTAL_MSBFS_TEST_SRCS + "${CMAKE_CURRENT_SOURCE_DIR}/experimental/ms_bfs_test.cpp") + +ConfigureTest(EXPERIMENTAL_MSBFS_TEST "${EXPERIMENTAL_MSBFS_TEST_SRCS}") + + ################################################################################################### # - Experimental SSSP tests ----------------------------------------------------------------------- diff --git a/cpp/tests/community/egonet_test.cu b/cpp/tests/community/egonet_test.cu index a9224b42bc1..e7fea43be42 100644 --- a/cpp/tests/community/egonet_test.cu +++ b/cpp/tests/community/egonet_test.cu @@ -182,150 +182,141 @@ INSTANTIATE_TEST_CASE_P( // For perf analysis /* INSTANTIATE_TEST_CASE_P( -simple_test, -Tests_InducedEgo, -::testing::Values( -InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 1, false), -InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 2, false), -InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 3, false), -InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 4, false), -InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 5, false), -InducedEgo_Usecase( -"test/datasets/soc-LiveJournal1.mtx", std::vector{363617}, 2, false), -InducedEgo_Usecase( -"test/datasets/soc-LiveJournal1.mtx", -std::vector{ - 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755}, - 2, - false), - InducedEgo_Usecase( - "test/datasets/soc-LiveJournal1.mtx", - std::vector{ - 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755, - 2536834, 3070144, 3888415, 3131712, 2382526, 1040771, 2631543, 4607218, 4465829, - 3341686, 2772973, 2611175, 4526129, 2624421, 1220593, 2593137, 3270705, 1503899, - 1213033, 4840102, 4529036, 3421116, 4264831, 4089751, 4272322, 3486998, 2830318, - 320953, 2388331, 520808, 3023094, 1600294, 3631119, 1716614, 4829213, 1175844, 960680, - 847662, 3277365, 3957318, 3455123, 2454259, 670953, 4465677, 1027332, 2560721, 89061, - 1163406, 3109528, 3221856, 4714426, 2382774, 37828, 4433616, 3283229, 591911, - 4200188, 442522, 872207, 2437601, 741003, 266241, 914618, 3626195, 2021080, - 4679624, 777476, 2527796, 1114017, 640142, 49259, 4069879, 3869098, 1105040, - 4707804, 3208582, 3325885, 1450601, 4072548, 2037062, 2029646, 4575891, 1488598, 79105, - 4827273, 3795434, 4647518, 4733397, 3980718, 1184627}, - 2, - false), - InducedEgo_Usecase( - "test/datasets/soc-LiveJournal1.mtx", - std::vector{ - 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755, - 2536834, 3070144, 3888415, 3131712, 2382526, 1040771, 2631543, 4607218, 4465829, - 3341686, 2772973, 2611175, 4526129, 2624421, 1220593, 2593137, 3270705, 1503899, - 1213033, 4840102, 4529036, 3421116, 4264831, 4089751, 4272322, 3486998, 2830318, - 320953, 2388331, 520808, 3023094, 1600294, 3631119, 1716614, 4829213, 1175844, 960680, - 847662, 3277365, 3957318, 3455123, 2454259, 670953, 4465677, 1027332, 2560721, 89061, - 1163406, 3109528, 3221856, 4714426, 2382774, 37828, 4433616, 3283229, 591911, - 4200188, 442522, 872207, 2437601, 741003, 266241, 914618, 3626195, 2021080, - 4679624, 777476, 2527796, 1114017, 640142, 49259, 4069879, 3869098, 1105040, - 4707804, 3208582, 3325885, 1450601, 4072548, 2037062, 2029646, 4575891, 1488598, 79105, - 4827273, 3795434, 4647518, 4733397, 3980718, 1184627, 984983, 3114832, 1967741, - 1599818, 144593, 2698770, 2889449, 2495550, 1053813, 1193622, 686026, 3989015, - 2040719, 4693428, 3190376, 2926728, 3399030, 1664419, 662429, 4526841, 2186957, - 3752558, 2440046, 2930226, 3633006, 4058166, 3137060, 3499296, 2126343, 148971, - 2199672, 275811, 2813976, 2274536, 1189239, 1335942, 2465624, 2596042, 829684, 193400, - 2682845, 3691697, 4022437, 4051170, 4195175, 2876420, 3984220, 2174475, 326134, - 2606530, 2493046, 4706121, 1498980, 4576225, 1271339, 44832, 1875673, 4664940, - 134931, 736397, 4333554, 2751031, 2163610, 2879676, 3174153, 3317403, 2052464, - 1881883, 4757859, 3596257, 2358088, 2578758, 447504, 590720, 1717038, 1869795, - 1133885, 3027521, 840312, 2818881, 3654321, 2730947, 353585, 1134903, 2223378, - 1508824, 3662521, 1363776, 2712071, 288441, 1204581, 3502242, 4645567, 2767267, - 1514366, 3956099, 1422145, 1216608, 2253360, 189132, 4238225, 1345783, 451571, 1599442, - 3237284, 4711405, 929446, 1857675, 150759, 1277633, 761210, 138628, 1026833, - 2599544, 2464737, 989203, 3399615, 2144292, 216142, 637312, 2044964, 716256, 1660632, - 1762919, 4784357, 2213415, 2764769, 291806, 609772, 3264819, 1870953, 1516385, - 235647, 1045474, 2664957, 819095, 1824119, 4045271, 4448109, 1676788, 4285177, - 1580502, 3546548, 2771971, 3927086, 1339779, 3156204, 1730998, 1172522, 2433024, - 4533449, 479930, 2010695, 672994, 3542039, 3176455, 26352, 2137735, 866910, - 4410835, 2623982, 3603159, 2555625, 2765653, 267865, 2015523, 1009052, 4713994, - 1600667, 2176195, 3179631, 4570390, 2018424, 3356384, 1784287, 894861, 3622099, - 1647273, 3044136, 950354, 1491760, 3416929, 3757300, 2244912, 4129215, 1600848, - 3867343, 72329, 919189, 992521, 3445975, 4712557, 4680974, 188419, 2612093, - 1991268, 3566207, 2281468, 3859078, 2492806, 3398628, 763441, 2679107, 2554420, - 2130132, 4664374, 1182901, 3890770, 4714667, 4209303, 4013060, 3617653, 2040022, - 3296519, 4190671, 1693353, 2678411, 3788834, 2781815, 191965, 1083926, 503974, 3529226, - 1650522, 1900976, 542080, 3423929, 3418905, 878165, 4701703, 3022790, 4316365, 76365, - 4053672, 1358185, 3830478, 4445661, 3210024, 1895915, 4541133, 2938808, 562788, - 3920065, 1458776, 4052046, 2967475, 1092809, 3203538, 159626, 3399464, 214467, - 3343982, 1811854, 3189045, 4272117, 4701563, 424807, 4341116, 760545, 4674683, - 1538018, 386762, 194237, 2162719, 1694433, 943728, 2389036, 2196653, 3085571, - 1513424, 3689413, 3278747, 4197291, 3324063, 3651090, 1737936, 2768803, 2768889, - 3108096, 4311775, 3569480, 886705, 733256, 2477493, 1735412, 2960895, 1983781, - 1861797, 3566460, 4537673, 1164093, 3499764, 4553071, 3518985, 847658, 918948, - 2922351, 1056144, 652895, 1013195, 780505, 1702928, 3562838, 1432719, 2405207, - 1054920, 641647, 2240939, 3617702, 383165, 652641, 879593, 1810739, 2096385, - 4497865, 4768530, 1743968, 3582014, 1025009, 3002122, 2422190, 527647, 1251821, - 2571153, 4095874, 3705333, 3637407, 1385567, 4043855, 4041930, 2433139, 1710383, - 1127734, 4362316, 711588, 817839, 3214775, 910077, 1313768, 2382229, 16864, 2081770, - 3095420, 3195272, 548711, 2259860, 1167323, 2435974, 425238, 2085179, 2630042, - 2632881, 2867923, 3703565, 1037695, 226617, 4379130, 1541468, 3581937, 605965, - 1137674, 4655221, 4769963, 1394370, 4425315, 2990132, 2364485, 1561137, 2713384, - 481509, 2900382, 934766, 2986774, 1767669, 298593, 2502539, 139296, 3794229, - 4002180, 4718138, 2909238, 423691, 3023810, 2784924, 2760160, 1971980, 316683, - 3828090, 3253691, 4839313, 1203624, 584938, 3901482, 1747543, 1572737, 3533226, - 774708, 1691195, 1037110, 1557763, 225120, 4424243, 3524086, 1717663, 4332507, - 3513592, 4274932, 1232118, 873498, 1416042, 2488925, 111391, 4704545, 4492545, - 445317, 1584812, 2187737, 2471948, 3731678, 219255, 2282627, 2589971, 2372185, - 4609096, 3673961, 2524410, 12823, 2437155, 3015974, 4188352, 3184084, 3690756, - 1222341, 1278376, 3652030, 4162647, 326548, 3930062, 3926100, 1551222, 2722165, - 4526695, 3997534, 4815513, 3139056, 2547644, 3028915, 4149092, 3656554, 2691582, - 2676699, 1878842, 260174, 3129900, 4379993, 182347, 2189338, 3783616, 2616666, - 2596952, 243007, 4179282, 2730, 1939894, 2332032, 3335636, 182332, 3112260, - 2174584, 587481, 4527368, 3154106, 3403059, 673206, 2150292, 446521, 1600204, - 4819428, 2591357, 48490, 2917012, 2285923, 1072926, 2824281, 4364250, 956033, 311938, - 37251, 3729300, 2726300, 644966, 1623020, 1419070, 4646747, 2417222, 2680238, - 2561083, 1793801, 2349366, 339747, 611366, 4684147, 4356907, 1277161, 4510381, - 3218352, 4161658, 3200733, 1172372, 3997786, 3169266, 3353418, 2248955, 2875885, - 2365369, 498208, 2968066, 2681505, 2059048, 2097106, 3607540, 1121504, 2016789, - 1762605, 3138431, 866081, 3705757, 3833066, 2599788, 760816, 4046672, 1544367, - 2983906, 4842911, 209599, 1250954, 3333704, 561212, 4674336, 2831841, 3690724, - 2929360, 4830834, 1177524, 2487687, 3525137, 875283, 651241, 2110742, 1296646, - 1543739, 4349417, 2384725, 1931751, 1519208, 1520034, 3385008, 3219962, 734912, 170230, - 1741419, 729913, 2860117, 2362381, 1199807, 2424230, 177824, 125948, 2722701, - 4687548, 1140771, 3232742, 4522020, 4376360, 1125603, 590312, 2481884, 138951, - 4086775, 615155, 3395781, 4587272, 283209, 568470, 4296185, 4344150, 2454321, - 2672602, 838828, 4051647, 1709120, 3074610, 693235, 4356087, 3018806, 239410, - 2431497, 691186, 766276, 4462126, 859155, 2370304, 1571808, 1938673, 1694955, - 3871296, 4245059, 3987376, 301524, 2512461, 3410437, 3300380, 684922, 4581995, - 3599557, 683515, 1850634, 3704678, 1937490, 2035591, 3718533, 2065879, 3160765, - 1467884, 1912241, 2501509, 3668572, 3390469, 2501150, 612319, 713633, 1976262, 135946, - 3641535, 632083, 13414, 4217765, 4137712, 2550250, 3281035, 4179598, 961045, - 2020694, 4380006, 1345936, 289162, 1359035, 770872, 4509911, 3947317, 4719693, - 248568, 2625660, 1237232, 2153208, 4814282, 1259954, 3677369, 861222, 2883506, - 3339149, 3998335, 491017, 1609022, 2648112, 742132, 649609, 4206953, 3131106, - 3504814, 3344486, 611721, 3215620, 2856233, 4447505, 1949222, 1868345, 712710, 6966, - 4730666, 3181872, 2972889, 3038521, 3525444, 4385208, 1845613, 1124187, 2030476, - 4468651, 2478792, 3473580, 3783357, 1852991, 1648485, 871319, 1670723, 4458328, - 3218600, 1811100, 3443356, 2233873, 3035207, 2548692, 3337891, 3773674, 1552957, - 4782811, 3144712, 3523466, 1491315, 3955852, 1838410, 3164028, 1092543, 776459, - 2959379, 2541744, 4064418, 3908320, 2854145, 3960709, 1348188, 977678, 853619, - 1304291, 2848702, 1657913, 1319826, 3322665, 788037, 2913686, 4471279, 1766285, 348304, - 56570, 1892118, 4017244, 401006, 3524539, 4310134, 1624693, 4081113, 957511, 849400, - 129975, 2616130, 378537, 1556787, 3916162, 1039980, 4407778, 2027690, 4213675, - 839863, 683134, 75805, 2493150, 4215796, 81587, 751845, 1255588, 1947964, - 1950470, 859401, 3077088, 3931110, 2316256, 1523761, 4527477, 4237511, 1123513, - 4209796, 3584772, 4250563, 2091754, 1618766, 2139944, 4525352, 382159, 2955887, 41760, - 2313998, 496912, 3791570, 3904792, 3613654, 873959, 127076, 2537797, 2458107, - 4543265, 3661909, 26828, 271816, 17854, 2461269, 1776042, 1573899, 3409957, - 4335712, 4534313, 3392751, 1230124, 2159031, 4444015, 3373087, 3848014, 2026600, - 1382747, 3537242, 4536743, 4714155, 3788371, 3570849, 173741, 211962, 4377778, - 119369, 2856973, 2945854, 1508054, 4503932, 3141566, 1842177, 3448683, 3384614, - 2886508, 1573965, 990618, 3053734, 2918742, 4508753, 1032149, 60943, 4291620, - 722607, 2883224, 169359, 4356585, 3725543, 3678729, 341673, 3592828, 4077251, - 3382936, 3885685, 4630994, 1286698, 4449616, 1138430, 3113385, 4660578, 2539973, - 4562286, 4085089, 494737, 3967610, 2130702, 1823755, 1369324, 3796951, 956299, 141730, - 935144, 4381893, 4412545, 1382250, 3024476, 2364546, 3396164, 3573511, 314081, 577688, - 4154135, 1567018, 4047761, 2446220, 1148833, 4842497, 3967186, 1175290, 3749667, - 1209593, 3295627, 3169065, 2460328, 1838486, 1436923, 2843887, 3676426, 2079145, - 2975635, 535071, 4287509, 3281107, 39606, 3115500, 3204573, 722131, 3124073}, -2, -false)));*/ + simple_test, + Tests_InducedEgo, + ::testing::Values( + InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 1, false), + InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 2, false), + InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 3, false), + InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 4, false), + InducedEgo_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{0}, 5, false), + InducedEgo_Usecase( + "test/datasets/soc-LiveJournal1.mtx", std::vector{363617}, 2, false), + InducedEgo_Usecase( + "test/datasets/soc-LiveJournal1.mtx", + std::vector{ + 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755}, + 2, + false), + InducedEgo_Usecase( + "test/datasets/soc-LiveJournal1.mtx", + std::vector{ + 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755, + 2536834, 3070144, 3888415, 3131712, 2382526, 1040771, 2631543, 4607218, 4465829, 3341686, + 2772973, 2611175, 4526129, 2624421, 1220593, 2593137, 3270705, 1503899, 1213033, 4840102, + 4529036, 3421116, 4264831, 4089751, 4272322, 3486998, 2830318, 320953, 2388331, 520808, + 3023094, 1600294, 3631119, 1716614, 4829213, 1175844, 960680, 847662, 3277365, 3957318, + 3455123, 2454259, 670953, 4465677, 1027332, 2560721, 89061, 1163406, 3109528, 3221856, + 4714426, 2382774, 37828, 4433616, 3283229, 591911, 4200188, 442522, 872207, 2437601, + 741003, 266241, 914618, 3626195, 2021080, 4679624, 777476, 2527796, 1114017, 640142, + 49259, 4069879, 3869098, 1105040, 4707804, 3208582, 3325885, 1450601, 4072548, 2037062, + 2029646, 4575891, 1488598, 79105, 4827273, 3795434, 4647518, 4733397, 3980718, 1184627}, + 2, + false), + InducedEgo_Usecase( + "test/datasets/soc-LiveJournal1.mtx", + std::vector{ + 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755, + 2536834, 3070144, 3888415, 3131712, 2382526, 1040771, 2631543, 4607218, 4465829, 3341686, + 2772973, 2611175, 4526129, 2624421, 1220593, 2593137, 3270705, 1503899, 1213033, 4840102, + 4529036, 3421116, 4264831, 4089751, 4272322, 3486998, 2830318, 320953, 2388331, 520808, + 3023094, 1600294, 3631119, 1716614, 4829213, 1175844, 960680, 847662, 3277365, 3957318, + 3455123, 2454259, 670953, 4465677, 1027332, 2560721, 89061, 1163406, 3109528, 3221856, + 4714426, 2382774, 37828, 4433616, 3283229, 591911, 4200188, 442522, 872207, 2437601, + 741003, 266241, 914618, 3626195, 2021080, 4679624, 777476, 2527796, 1114017, 640142, + 49259, 4069879, 3869098, 1105040, 4707804, 3208582, 3325885, 1450601, 4072548, 2037062, + 2029646, 4575891, 1488598, 79105, 4827273, 3795434, 4647518, 4733397, 3980718, 1184627, + 984983, 3114832, 1967741, 1599818, 144593, 2698770, 2889449, 2495550, 1053813, 1193622, + 686026, 3989015, 2040719, 4693428, 3190376, 2926728, 3399030, 1664419, 662429, 4526841, + 2186957, 3752558, 2440046, 2930226, 3633006, 4058166, 3137060, 3499296, 2126343, 148971, + 2199672, 275811, 2813976, 2274536, 1189239, 1335942, 2465624, 2596042, 829684, 193400, + 2682845, 3691697, 4022437, 4051170, 4195175, 2876420, 3984220, 2174475, 326134, 2606530, + 2493046, 4706121, 1498980, 4576225, 1271339, 44832, 1875673, 4664940, 134931, 736397, + 4333554, 2751031, 2163610, 2879676, 3174153, 3317403, 2052464, 1881883, 4757859, 3596257, + 2358088, 2578758, 447504, 590720, 1717038, 1869795, 1133885, 3027521, 840312, 2818881, + 3654321, 2730947, 353585, 1134903, 2223378, 1508824, 3662521, 1363776, 2712071, 288441, + 1204581, 3502242, 4645567, 2767267, 1514366, 3956099, 1422145, 1216608, 2253360, 189132, + 4238225, 1345783, 451571, 1599442, 3237284, 4711405, 929446, 1857675, 150759, 1277633, + 761210, 138628, 1026833, 2599544, 2464737, 989203, 3399615, 2144292, 216142, 637312, + 2044964, 716256, 1660632, 1762919, 4784357, 2213415, 2764769, 291806, 609772, 3264819, + 1870953, 1516385, 235647, 1045474, 2664957, 819095, 1824119, 4045271, 4448109, 1676788, + 4285177, 1580502, 3546548, 2771971, 3927086, 1339779, 3156204, 1730998, 1172522, 2433024, + 4533449, 479930, 2010695, 672994, 3542039, 3176455, 26352, 2137735, 866910, 4410835, + 2623982, 3603159, 2555625, 2765653, 267865, 2015523, 1009052, 4713994, 1600667, 2176195, + 3179631, 4570390, 2018424, 3356384, 1784287, 894861, 3622099, 1647273, 3044136, 950354, + 1491760, 3416929, 3757300, 2244912, 4129215, 1600848, 3867343, 72329, 919189, 992521, + 3445975, 4712557, 4680974, 188419, 2612093, 1991268, 3566207, 2281468, 3859078, 2492806, + 3398628, 763441, 2679107, 2554420, 2130132, 4664374, 1182901, 3890770, 4714667, 4209303, + 4013060, 3617653, 2040022, 3296519, 4190671, 1693353, 2678411, 3788834, 2781815, 191965, + 1083926, 503974, 3529226, 1650522, 1900976, 542080, 3423929, 3418905, 878165, 4701703, + 3022790, 4316365, 76365, 4053672, 1358185, 3830478, 4445661, 3210024, 1895915, 4541133, + 2938808, 562788, 3920065, 1458776, 4052046, 2967475, 1092809, 3203538, 159626, 3399464, + 214467, 3343982, 1811854, 3189045, 4272117, 4701563, 424807, 4341116, 760545, 4674683, + 1538018, 386762, 194237, 2162719, 1694433, 943728, 2389036, 2196653, 3085571, 1513424, + 3689413, 3278747, 4197291, 3324063, 3651090, 1737936, 2768803, 2768889, 3108096, 4311775, + 3569480, 886705, 733256, 2477493, 1735412, 2960895, 1983781, 1861797, 3566460, 4537673, + 1164093, 3499764, 4553071, 3518985, 847658, 918948, 2922351, 1056144, 652895, 1013195, + 780505, 1702928, 3562838, 1432719, 2405207, 1054920, 641647, 2240939, 3617702, 383165, + 652641, 879593, 1810739, 2096385, 4497865, 4768530, 1743968, 3582014, 1025009, 3002122, + 2422190, 527647, 1251821, 2571153, 4095874, 3705333, 3637407, 1385567, 4043855, 4041930, + 2433139, 1710383, 1127734, 4362316, 711588, 817839, 3214775, 910077, 1313768, 2382229, + 16864, 2081770, 3095420, 3195272, 548711, 2259860, 1167323, 2435974, 425238, 2085179, + 2630042, 2632881, 2867923, 3703565, 1037695, 226617, 4379130, 1541468, 3581937, 605965, + 1137674, 4655221, 4769963, 1394370, 4425315, 2990132, 2364485, 1561137, 2713384, 481509, + 2900382, 934766, 2986774, 1767669, 298593, 2502539, 139296, 3794229, 4002180, 4718138, + 2909238, 423691, 3023810, 2784924, 2760160, 1971980, 316683, 3828090, 3253691, 4839313, + 1203624, 584938, 3901482, 1747543, 1572737, 3533226, 774708, 1691195, 1037110, 1557763, + 225120, 4424243, 3524086, 1717663, 4332507, 3513592, 4274932, 1232118, 873498, 1416042, + 2488925, 111391, 4704545, 4492545, 445317, 1584812, 2187737, 2471948, 3731678, 219255, + 2282627, 2589971, 2372185, 4609096, 3673961, 2524410, 12823, 2437155, 3015974, 4188352, + 3184084, 3690756, 1222341, 1278376, 3652030, 4162647, 326548, 3930062, 3926100, 1551222, + 2722165, 4526695, 3997534, 4815513, 3139056, 2547644, 3028915, 4149092, 3656554, 2691582, + 2676699, 1878842, 260174, 3129900, 4379993, 182347, 2189338, 3783616, 2616666, 2596952, + 243007, 4179282, 2730, 1939894, 2332032, 3335636, 182332, 3112260, 2174584, 587481, + 4527368, 3154106, 3403059, 673206, 2150292, 446521, 1600204, 4819428, 2591357, 48490, + 2917012, 2285923, 1072926, 2824281, 4364250, 956033, 311938, 37251, 3729300, 2726300, + 644966, 1623020, 1419070, 4646747, 2417222, 2680238, 2561083, 1793801, 2349366, 339747, + 611366, 4684147, 4356907, 1277161, 4510381, 3218352, 4161658, 3200733, 1172372, 3997786, + 3169266, 3353418, 2248955, 2875885, 2365369, 498208, 2968066, 2681505, 2059048, 2097106, + 3607540, 1121504, 2016789, 1762605, 3138431, 866081, 3705757, 3833066, 2599788, 760816, + 4046672, 1544367, 2983906, 4842911, 209599, 1250954, 3333704, 561212, 4674336, 2831841, + 3690724, 2929360, 4830834, 1177524, 2487687, 3525137, 875283, 651241, 2110742, 1296646, + 1543739, 4349417, 2384725, 1931751, 1519208, 1520034, 3385008, 3219962, 734912, 170230, + 1741419, 729913, 2860117, 2362381, 1199807, 2424230, 177824, 125948, 2722701, 4687548, + 1140771, 3232742, 4522020, 4376360, 1125603, 590312, 2481884, 138951, 4086775, 615155, + 3395781, 4587272, 283209, 568470, 4296185, 4344150, 2454321, 2672602, 838828, 4051647, + 1709120, 3074610, 693235, 4356087, 3018806, 239410, 2431497, 691186, 766276, 4462126, + 859155, 2370304, 1571808, 1938673, 1694955, 3871296, 4245059, 3987376, 301524, 2512461, + 3410437, 3300380, 684922, 4581995, 3599557, 683515, 1850634, 3704678, 1937490, 2035591, + 3718533, 2065879, 3160765, 1467884, 1912241, 2501509, 3668572, 3390469, 2501150, 612319, + 713633, 1976262, 135946, 3641535, 632083, 13414, 4217765, 4137712, 2550250, 3281035, + 4179598, 961045, 2020694, 4380006, 1345936, 289162, 1359035, 770872, 4509911, 3947317, + 4719693, 248568, 2625660, 1237232, 2153208, 4814282, 1259954, 3677369, 861222, 2883506, + 3339149, 3998335, 491017, 1609022, 2648112, 742132, 649609, 4206953, 3131106, 3504814, + 3344486, 611721, 3215620, 2856233, 4447505, 1949222, 1868345, 712710, 6966, 4730666, + 3181872, 2972889, 3038521, 3525444, 4385208, 1845613, 1124187, 2030476, 4468651, 2478792, + 3473580, 3783357, 1852991, 1648485, 871319, 1670723, 4458328, 3218600, 1811100, 3443356, + 2233873, 3035207, 2548692, 3337891, 3773674, 1552957, 4782811, 3144712, 3523466, 1491315, + 3955852, 1838410, 3164028, 1092543, 776459, 2959379, 2541744, 4064418, 3908320, 2854145, + 3960709, 1348188, 977678, 853619, 1304291, 2848702, 1657913, 1319826, 3322665, 788037, + 2913686, 4471279, 1766285, 348304, 56570, 1892118, 4017244, 401006, 3524539, 4310134, + 1624693, 4081113, 957511, 849400, 129975, 2616130, 378537, 1556787, 3916162, 1039980, + 4407778, 2027690, 4213675, 839863, 683134, 75805, 2493150, 4215796, 81587, 751845, + 1255588, 1947964, 1950470, 859401, 3077088, 3931110, 2316256, 1523761, 4527477, 4237511, + 1123513, 4209796, 3584772, 4250563, 2091754, 1618766, 2139944, 4525352, 382159, 2955887, + 41760, 2313998, 496912, 3791570, 3904792, 3613654, 873959, 127076, 2537797, 2458107, + 4543265, 3661909, 26828, 271816, 17854, 2461269, 1776042, 1573899, 3409957, 4335712, + 4534313, 3392751, 1230124, 2159031, 4444015, 3373087, 3848014, 2026600, 1382747, 3537242, + 4536743, 4714155, 3788371, 3570849, 173741, 211962, 4377778, 119369, 2856973, 2945854, + 1508054, 4503932, 3141566, 1842177, 3448683, 3384614, 2886508, 1573965, 990618, 3053734, + 2918742, 4508753, 1032149, 60943, 4291620, 722607, 2883224, 169359, 4356585, 3725543, + 3678729, 341673, 3592828, 4077251, 3382936, 3885685, 4630994, 1286698, 4449616, 1138430, + 3113385, 4660578, 2539973, 4562286, 4085089, 494737, 3967610, 2130702, 1823755, 1369324, + 3796951, 956299, 141730, 935144, 4381893, 4412545, 1382250, 3024476, 2364546, 3396164, + 3573511, 314081, 577688, 4154135, 1567018, 4047761, 2446220, 1148833, 4842497, 3967186, + 1175290, 3749667, 1209593, 3295627, 3169065, 2460328, 1838486, 1436923, 2843887, 3676426, + 2079145, 2975635, 535071, 4287509, 3281107, 39606, 3115500, 3204573, 722131, 3124073}, + 2, + false))); +*/ CUGRAPH_TEST_PROGRAM_MAIN() diff --git a/cpp/tests/experimental/ms_bfs_test.cpp b/cpp/tests/experimental/ms_bfs_test.cpp new file mode 100644 index 00000000000..264382c22a3 --- /dev/null +++ b/cpp/tests/experimental/ms_bfs_test.cpp @@ -0,0 +1,301 @@ +/* + * Copyright (c) 2021, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include +#include +#include + +#include +#include +#include +#include + +#include +#include +#include +#include + +#include + +#include +#include +#include +#include +#include +#include +#include + +typedef struct MsBfs_Usecase_t { + std::string graph_file_full_path{}; + std::vector sources{}; + int32_t radius; + bool test_weighted{false}; + + MsBfs_Usecase_t(std::string const& graph_file_path, + std::vector const& sources, + int32_t radius, + bool test_weighted) + : sources(sources), radius(radius), test_weighted(test_weighted) + { + if ((graph_file_path.length() > 0) && (graph_file_path[0] != '/')) { + graph_file_full_path = cugraph::test::get_rapids_dataset_root_dir() + "/" + graph_file_path; + } else { + graph_file_full_path = graph_file_path; + } + }; +} MsBfs_Usecase; + +class Tests_MsBfs : public ::testing::TestWithParam { + public: + Tests_MsBfs() {} + static void SetupTestCase() {} + static void TearDownTestCase() {} + + virtual void SetUp() {} + virtual void TearDown() {} + + template + void run_current_test(MsBfs_Usecase const& configuration) + { + auto n_seeds = configuration.sources.size(); + int n_streams = std::min(n_seeds, static_cast(128)); + raft::handle_t handle(n_streams); + + cugraph::experimental::graph_t graph( + handle); + std::tie(graph, std::ignore) = cugraph::test:: + read_graph_from_matrix_market_file( + handle, configuration.graph_file_full_path, configuration.test_weighted, false); + auto graph_view = graph.view(); + // Streams will allocate concurrently later + std::vector> d_distances{}; + std::vector> d_predecessors{}; + + d_distances.reserve(n_seeds); + d_predecessors.reserve(n_seeds); + for (vertex_t i = 0; i < n_seeds; i++) { + // Allocations and operations are attached to the worker stream + rmm::device_uvector tmp_distances(graph_view.get_number_of_vertices(), + handle.get_internal_stream_view(i)); + rmm::device_uvector tmp_predecessors(graph_view.get_number_of_vertices(), + handle.get_internal_stream_view(i)); + + d_distances.push_back(std::move(tmp_distances)); + d_predecessors.push_back(std::move(tmp_predecessors)); + } + + std::vector radius(n_seeds); + std::generate(radius.begin(), radius.end(), [n = 0]() mutable { return (n++ % 12 + 1); }); + + // warm up + cugraph::experimental::bfs(handle, + graph_view, + d_distances[0].begin(), + d_predecessors[0].begin(), + static_cast(configuration.sources[0]), + false, + radius[0]); + + // one by one + HighResTimer hr_timer; + hr_timer.start("bfs"); + cudaProfilerStart(); + for (vertex_t i = 0; i < n_seeds; i++) { + cugraph::experimental::bfs(handle, + graph_view, + d_distances[i].begin(), + d_predecessors[i].begin(), + static_cast(configuration.sources[i]), + false, + radius[i]); + } + cudaProfilerStop(); + hr_timer.stop(); + hr_timer.display(std::cout); + + // concurrent + hr_timer.start("bfs"); + cudaProfilerStart(); +#pragma omp parallel for + for (vertex_t i = 0; i < n_seeds; i++) { + raft::handle_t light_handle(handle, i); + auto worker_stream_view = light_handle.get_stream_view(); + cugraph::experimental::bfs(light_handle, + graph_view, + d_distances[i].begin(), + d_predecessors[i].begin(), + static_cast(configuration.sources[i]), + false, + radius[i]); + } + + cudaProfilerStop(); + hr_timer.stop(); + hr_timer.display(std::cout); + } +}; + +TEST_P(Tests_MsBfs, DISABLED_CheckInt32Int32FloatUntransposed) +{ + run_current_test(GetParam()); +} +/* +INSTANTIATE_TEST_CASE_P( + simple_test, + Tests_MsBfs, + ::testing::Values( + MsBfs_Usecase("test/datasets/karate.mtx", std::vector{0}, 1, false), + MsBfs_Usecase("test/datasets/karate.mtx", std::vector{0}, 2, false), + MsBfs_Usecase("test/datasets/karate.mtx", std::vector{1}, 3, false), + MsBfs_Usecase("test/datasets/karate.mtx", std::vector{10, 0, 5}, 2, false), + MsBfs_Usecase("test/datasets/karate.mtx", std::vector{9, 3, 10}, 2, false), + MsBfs_Usecase( + "test/datasets/karate.mtx", std::vector{5, 9, 3, 10, 12, 13}, 2, true))); +*/ +// For perf analysis + +INSTANTIATE_TEST_CASE_P( + simple_test, + Tests_MsBfs, + ::testing::Values( + MsBfs_Usecase("test/datasets/soc-LiveJournal1.mtx", std::vector{363617}, 2, false), + MsBfs_Usecase( + "test/datasets/soc-LiveJournal1.mtx", + std::vector{ + 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755}, + 2, + false), + MsBfs_Usecase( + "test/datasets/soc-LiveJournal1.mtx", + std::vector{ + 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755, + 2536834, 3070144, 3888415, 3131712, 2382526, 1040771, 2631543, 4607218, 4465829, 3341686, + 2772973, 2611175, 4526129, 2624421, 1220593, 2593137, 3270705, 1503899, 1213033, 4840102, + 4529036, 3421116, 4264831, 4089751, 4272322, 3486998, 2830318, 320953, 2388331, 520808, + 3023094, 1600294, 3631119, 1716614, 4829213, 1175844, 960680, 847662, 3277365, 3957318, + 3455123, 2454259, 670953, 4465677, 1027332, 2560721, 89061, 1163406, 3109528, 3221856, + 4714426, 2382774, 37828, 4433616, 3283229, 591911, 4200188, 442522, 872207, 2437601, + 741003, 266241, 914618, 3626195, 2021080, 4679624, 777476, 2527796, 1114017, 640142, + 49259, 4069879, 3869098, 1105040, 4707804, 3208582, 3325885, 1450601, 4072548, 2037062, + 2029646, 4575891, 1488598, 79105, 4827273, 3795434, 4647518, 4733397, 3980718, 1184627}, + 2, + false), + MsBfs_Usecase( + "test/datasets/soc-LiveJournal1.mtx", + std::vector{ + 363617, 722214, 2337449, 2510183, 2513389, 225853, 2035807, 3836330, 1865496, 28755, + 2536834, 3070144, 3888415, 3131712, 2382526, 1040771, 2631543, 4607218, 4465829, 3341686, + 2772973, 2611175, 4526129, 2624421, 1220593, 2593137, 3270705, 1503899, 1213033, 4840102, + 4529036, 3421116, 4264831, 4089751, 4272322, 3486998, 2830318, 320953, 2388331, 520808, + 3023094, 1600294, 3631119, 1716614, 4829213, 1175844, 960680, 847662, 3277365, 3957318, + 3455123, 2454259, 670953, 4465677, 1027332, 2560721, 89061, 1163406, 3109528, 3221856, + 4714426, 2382774, 37828, 4433616, 3283229, 591911, 4200188, 442522, 872207, 2437601, + 741003, 266241, 914618, 3626195, 2021080, 4679624, 777476, 2527796, 1114017, 640142, + 49259, 4069879, 3869098, 1105040, 4707804, 3208582, 3325885, 1450601, 4072548, 2037062, + 2029646, 4575891, 1488598, 79105, 4827273, 3795434, 4647518, 4733397, 3980718, 1184627, + 984983, 3114832, 1967741, 1599818, 144593, 2698770, 2889449, 2495550, 1053813, 1193622, + 686026, 3989015, 2040719, 4693428, 3190376, 2926728, 3399030, 1664419, 662429, 4526841, + 2186957, 3752558, 2440046, 2930226, 3633006, 4058166, 3137060, 3499296, 2126343, 148971, + 2199672, 275811, 2813976, 2274536, 1189239, 1335942, 2465624, 2596042, 829684, 193400, + 2682845, 3691697, 4022437, 4051170, 4195175, 2876420, 3984220, 2174475, 326134, 2606530, + 2493046, 4706121, 1498980, 4576225, 1271339, 44832, 1875673, 4664940, 134931, 736397, + 4333554, 2751031, 2163610, 2879676, 3174153, 3317403, 2052464, 1881883, 4757859, 3596257, + 2358088, 2578758, 447504, 590720, 1717038, 1869795, 1133885, 3027521, 840312, 2818881, + 3654321, 2730947, 353585, 1134903, 2223378, 1508824, 3662521, 1363776, 2712071, 288441, + 1204581, 3502242, 4645567, 2767267, 1514366, 3956099, 1422145, 1216608, 2253360, 189132, + 4238225, 1345783, 451571, 1599442, 3237284, 4711405, 929446, 1857675, 150759, 1277633, + 761210, 138628, 1026833, 2599544, 2464737, 989203, 3399615, 2144292, 216142, 637312, + 2044964, 716256, 1660632, 1762919, 4784357, 2213415, 2764769, 291806, 609772, 3264819, + 1870953, 1516385, 235647, 1045474, 2664957, 819095, 1824119, 4045271, 4448109, 1676788, + 4285177, 1580502, 3546548, 2771971, 3927086, 1339779, 3156204, 1730998, 1172522, 2433024, + 4533449, 479930, 2010695, 672994, 3542039, 3176455, 26352, 2137735, 866910, 4410835, + 2623982, 3603159, 2555625, 2765653, 267865, 2015523, 1009052, 4713994, 1600667, 2176195, + 3179631, 4570390, 2018424, 3356384, 1784287, 894861, 3622099, 1647273, 3044136, 950354, + 1491760, 3416929, 3757300, 2244912, 4129215, 1600848, 3867343, 72329, 919189, 992521, + 3445975, 4712557, 4680974, 188419, 2612093, 1991268, 3566207, 2281468, 3859078, 2492806, + 3398628, 763441, 2679107, 2554420, 2130132, 4664374, 1182901, 3890770, 4714667, 4209303, + 4013060, 3617653, 2040022, 3296519, 4190671, 1693353, 2678411, 3788834, 2781815, 191965, + 1083926, 503974, 3529226, 1650522, 1900976, 542080, 3423929, 3418905, 878165, 4701703, + 3022790, 4316365, 76365, 4053672, 1358185, 3830478, 4445661, 3210024, 1895915, 4541133, + 2938808, 562788, 3920065, 1458776, 4052046, 2967475, 1092809, 3203538, 159626, 3399464, + 214467, 3343982, 1811854, 3189045, 4272117, 4701563, 424807, 4341116, 760545, 4674683, + 1538018, 386762, 194237, 2162719, 1694433, 943728, 2389036, 2196653, 3085571, 1513424, + 3689413, 3278747, 4197291, 3324063, 3651090, 1737936, 2768803, 2768889, 3108096, 4311775, + 3569480, 886705, 733256, 2477493, 1735412, 2960895, 1983781, 1861797, 3566460, 4537673, + 1164093, 3499764, 4553071, 3518985, 847658, 918948, 2922351, 1056144, 652895, 1013195, + 780505, 1702928, 3562838, 1432719, 2405207, 1054920, 641647, 2240939, 3617702, 383165, + 652641, 879593, 1810739, 2096385, 4497865, 4768530, 1743968, 3582014, 1025009, 3002122, + 2422190, 527647, 1251821, 2571153, 4095874, 3705333, 3637407, 1385567, 4043855, 4041930, + 2433139, 1710383, 1127734, 4362316, 711588, 817839, 3214775, 910077, 1313768, 2382229, + 16864, 2081770, 3095420, 3195272, 548711, 2259860, 1167323, 2435974, 425238, 2085179, + 2630042, 2632881, 2867923, 3703565, 1037695, 226617, 4379130, 1541468, 3581937, 605965, + 1137674, 4655221, 4769963, 1394370, 4425315, 2990132, 2364485, 1561137, 2713384, 481509, + 2900382, 934766, 2986774, 1767669, 298593, 2502539, 139296, 3794229, 4002180, 4718138, + 2909238, 423691, 3023810, 2784924, 2760160, 1971980, 316683, 3828090, 3253691, 4839313, + 1203624, 584938, 3901482, 1747543, 1572737, 3533226, 774708, 1691195, 1037110, 1557763, + 225120, 4424243, 3524086, 1717663, 4332507, 3513592, 4274932, 1232118, 873498, 1416042, + 2488925, 111391, 4704545, 4492545, 445317, 1584812, 2187737, 2471948, 3731678, 219255, + 2282627, 2589971, 2372185, 4609096, 3673961, 2524410, 12823, 2437155, 3015974, 4188352, + 3184084, 3690756, 1222341, 1278376, 3652030, 4162647, 326548, 3930062, 3926100, 1551222, + 2722165, 4526695, 3997534, 4815513, 3139056, 2547644, 3028915, 4149092, 3656554, 2691582, + 2676699, 1878842, 260174, 3129900, 4379993, 182347, 2189338, 3783616, 2616666, 2596952, + 243007, 4179282, 2730, 1939894, 2332032, 3335636, 182332, 3112260, 2174584, 587481, + 4527368, 3154106, 3403059, 673206, 2150292, 446521, 1600204, 4819428, 2591357, 48490, + 2917012, 2285923, 1072926, 2824281, 4364250, 956033, 311938, 37251, 3729300, 2726300, + 644966, 1623020, 1419070, 4646747, 2417222, 2680238, 2561083, 1793801, 2349366, 339747, + 611366, 4684147, 4356907, 1277161, 4510381, 3218352, 4161658, 3200733, 1172372, 3997786, + 3169266, 3353418, 2248955, 2875885, 2365369, 498208, 2968066, 2681505, 2059048, 2097106, + 3607540, 1121504, 2016789, 1762605, 3138431, 866081, 3705757, 3833066, 2599788, 760816, + 4046672, 1544367, 2983906, 4842911, 209599, 1250954, 3333704, 561212, 4674336, 2831841, + 3690724, 2929360, 4830834, 1177524, 2487687, 3525137, 875283, 651241, 2110742, 1296646, + 1543739, 4349417, 2384725, 1931751, 1519208, 1520034, 3385008, 3219962, 734912, 170230, + 1741419, 729913, 2860117, 2362381, 1199807, 2424230, 177824, 125948, 2722701, 4687548, + 1140771, 3232742, 4522020, 4376360, 1125603, 590312, 2481884, 138951, 4086775, 615155, + 3395781, 4587272, 283209, 568470, 4296185, 4344150, 2454321, 2672602, 838828, 4051647, + 1709120, 3074610, 693235, 4356087, 3018806, 239410, 2431497, 691186, 766276, 4462126, + 859155, 2370304, 1571808, 1938673, 1694955, 3871296, 4245059, 3987376, 301524, 2512461, + 3410437, 3300380, 684922, 4581995, 3599557, 683515, 1850634, 3704678, 1937490, 2035591, + 3718533, 2065879, 3160765, 1467884, 1912241, 2501509, 3668572, 3390469, 2501150, 612319, + 713633, 1976262, 135946, 3641535, 632083, 13414, 4217765, 4137712, 2550250, 3281035, + 4179598, 961045, 2020694, 4380006, 1345936, 289162, 1359035, 770872, 4509911, 3947317, + 4719693, 248568, 2625660, 1237232, 2153208, 4814282, 1259954, 3677369, 861222, 2883506, + 3339149, 3998335, 491017, 1609022, 2648112, 742132, 649609, 4206953, 3131106, 3504814, + 3344486, 611721, 3215620, 2856233, 4447505, 1949222, 1868345, 712710, 6966, 4730666, + 3181872, 2972889, 3038521, 3525444, 4385208, 1845613, 1124187, 2030476, 4468651, 2478792, + 3473580, 3783357, 1852991, 1648485, 871319, 1670723, 4458328, 3218600, 1811100, 3443356, + 2233873, 3035207, 2548692, 3337891, 3773674, 1552957, 4782811, 3144712, 3523466, 1491315, + 3955852, 1838410, 3164028, 1092543, 776459, 2959379, 2541744, 4064418, 3908320, 2854145, + 3960709, 1348188, 977678, 853619, 1304291, 2848702, 1657913, 1319826, 3322665, 788037, + 2913686, 4471279, 1766285, 348304, 56570, 1892118, 4017244, 401006, 3524539, 4310134, + 1624693, 4081113, 957511, 849400, 129975, 2616130, 378537, 1556787, 3916162, 1039980, + 4407778, 2027690, 4213675, 839863, 683134, 75805, 2493150, 4215796, 81587, 751845, + 1255588, 1947964, 1950470, 859401, 3077088, 3931110, 2316256, 1523761, 4527477, 4237511, + 1123513, 4209796, 3584772, 4250563, 2091754, 1618766, 2139944, 4525352, 382159, 2955887, + 41760, 2313998, 496912, 3791570, 3904792, 3613654, 873959, 127076, 2537797, 2458107, + 4543265, 3661909, 26828, 271816, 17854, 2461269, 1776042, 1573899, 3409957, 4335712, + 4534313, 3392751, 1230124, 2159031, 4444015, 3373087, 3848014, 2026600, 1382747, 3537242, + 4536743, 4714155, 3788371, 3570849, 173741, 211962, 4377778, 119369, 2856973, 2945854, + 1508054, 4503932, 3141566, 1842177, 3448683, 3384614, 2886508, 1573965, 990618, 3053734, + 2918742, 4508753, 1032149, 60943, 4291620, 722607, 2883224, 169359, 4356585, 3725543, + 3678729, 341673, 3592828, 4077251, 3382936, 3885685, 4630994, 1286698, 4449616, 1138430, + 3113385, 4660578, 2539973, 4562286, 4085089, 494737, 3967610, 2130702, 1823755, 1369324, + 3796951, 956299, 141730, 935144, 4381893, 4412545, 1382250, 3024476, 2364546, 3396164, + 3573511, 314081, 577688, 4154135, 1567018, 4047761, 2446220, 1148833, 4842497, 3967186, + 1175290, 3749667, 1209593, 3295627, 3169065, 2460328, 1838486, 1436923, 2843887, 3676426, + 2079145, 2975635, 535071, 4287509, 3281107, 39606, 3115500, 3204573, 722131, 3124073}, + 2, + false))); +CUGRAPH_TEST_PROGRAM_MAIN() diff --git a/cpp/tests/experimental/streams.cu b/cpp/tests/experimental/streams.cu new file mode 100644 index 00000000000..c89ffe1e532 --- /dev/null +++ b/cpp/tests/experimental/streams.cu @@ -0,0 +1,44 @@ +/* + * Copyright (c) 2021, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ + +#include +#include +#include +#include +#include +#include +#include "gtest/gtest.h" +struct StreamTest : public ::testing::Test { +}; +TEST_F(StreamTest, basic_test) +{ + int n_streams = 4; + raft::handle_t handle(n_streams); + + const size_t intput_size = 4096; + +#pragma omp parallel for + for (int i = 0; i < n_streams; i++) { + rmm::device_uvector u(intput_size, handle.get_internal_stream_view(i)), + v(intput_size, handle.get_internal_stream_view(i)); + thrust::transform(rmm::exec_policy(handle.get_internal_stream_view(i)), + u.begin(), + u.end(), + v.begin(), + v.begin(), + 2 * thrust::placeholders::_1 + thrust::placeholders::_2); + } +} \ No newline at end of file diff --git a/python/cugraph/__init__.py b/python/cugraph/__init__.py index 8a847d1f1d4..11ba2d6ef96 100644 --- a/python/cugraph/__init__.py +++ b/python/cugraph/__init__.py @@ -82,7 +82,9 @@ shortest_path, filter_unreachable, shortest_path_length, - traveling_salesperson + traveling_salesperson, + concurrent_bfs, + multi_source_bfs, ) from cugraph.tree import minimum_spanning_tree, maximum_spanning_tree diff --git a/python/cugraph/tests/test_egonet.py b/python/cugraph/tests/test_egonet.py index 009fd1252f1..b259c2567dc 100644 --- a/python/cugraph/tests/test_egonet.py +++ b/python/cugraph/tests/test_egonet.py @@ -58,29 +58,6 @@ def test_ego_graph_nx(graph_file, seed, radius): @pytest.mark.parametrize("seeds", [[0, 5, 13]]) @pytest.mark.parametrize("radius", [1, 2, 3]) def test_batched_ego_graphs(graph_file, seeds, radius): - """ - Compute the induced subgraph of neighbors for each node in seeds - within a given radius. - Parameters - ---------- - G : cugraph.Graph, networkx.Graph, CuPy or SciPy sparse matrix - Graph or matrix object, which should contain the connectivity - information. Edge weights, if present, should be single or double - precision floating point values. - seeds : cudf.Series - Specifies the seeds of the induced egonet subgraphs - radius: integer, optional - Include all neighbors of distance<=radius from n. - - Returns - ------- - ego_edge_lists : cudf.DataFrame - GPU data frame containing all induced sources identifiers, - destination identifiers, edge weights - seeds_offsets: cudf.Series - Series containing the starting offset in the returned edge list - for each seed. - """ gc.collect() # Nx @@ -93,9 +70,8 @@ def test_batched_ego_graphs(graph_file, seeds, radius): df, offsets = cugraph.batched_ego_graphs(Gnx, seeds, radius=radius) for i in range(len(seeds)): ego_nx = nx.ego_graph(Gnx, seeds[i], radius=radius) - ego_df = df[offsets[i]:offsets[i+1]] - ego_cugraph = nx.from_pandas_edgelist(ego_df, - source="src", - target="dst", - edge_attr="weight") + ego_df = df[offsets[i]:offsets[i + 1]] + ego_cugraph = nx.from_pandas_edgelist( + ego_df, source="src", target="dst", edge_attr="weight" + ) assert nx.is_isomorphic(ego_nx, ego_cugraph) diff --git a/python/cugraph/traversal/__init__.py b/python/cugraph/traversal/__init__.py index 5944ebe0865..e74266d29fc 100644 --- a/python/cugraph/traversal/__init__.py +++ b/python/cugraph/traversal/__init__.py @@ -17,6 +17,8 @@ sssp, shortest_path, filter_unreachable, - shortest_path_length + shortest_path_length, ) from cugraph.traversal.traveling_salesperson import traveling_salesperson + +from cugraph.traversal.ms_bfs import concurrent_bfs, multi_source_bfs diff --git a/python/cugraph/traversal/ms_bfs.py b/python/cugraph/traversal/ms_bfs.py new file mode 100644 index 00000000000..e4b799e30e4 --- /dev/null +++ b/python/cugraph/traversal/ms_bfs.py @@ -0,0 +1,282 @@ +# Copyright (c) 2021, NVIDIA CORPORATION. +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +import numpy as np +import cudf + +# from cugraph.structure.graph import Graph, DiGraph +# from cugraph.utilities.utils import get_device_memory_info +import warnings + + +def _get_feasibility(G, sources, components=None, depth_limit=None): + """ + Evaluate the feasibility for breadth first traversal from multiple sources + in a graph. + + Parameters + ---------- + G : cugraph.Graph or cugraph.DiGraph + The adjacency list will be computed if not already present. + + sources : cudf.Series + Subset of vertices from which the traversals start. A BFS is run for + each source in the Series. + The size of the series should be at least one and cannot exceed + the size of the graph. + + depth_limit : Integer, optional, default=None + Limit the depth of the search. Terminates if no more vertices are + reachable within the distance of depth_limit + + components : cudf.DataFrame, optional, default=None + GPU Dataframe containing the component information. + Passing this information may impact the return type. + When no component information is passed BFS uses one component + behavior settings. + + components['vertex'] : cudf.Series + vertex IDs + components['color'] : cudf.Series + component IDs/color for vertices. + + Returns + ------- + mem_footprint : integer + Estimated memory foot print size in Bytes + """ + + # Fixme not implemented in RMM yet + # using 96GB upper bound for now + # mem = get_device_memory_info() + mem = 9.6e10 + n_sources = sources.size + V = G.number_of_vertices() + E = G.number_of_edges() + mean_component_sz = V + n_components = 1 + + # Retreive types + size_of_v = 4 + size_of_e = 4 + size_of_w = 0 + if G.adjlist.weights is not None: + if G.adjlist.weights.dtype is np.float64: + size_of_w = 8 + else: + size_of_w = 4 + if G.adjlist.offsets.dtype is np.float64: + size_of_v = 8 + if G.adjlist.indices.dtype is np.float64: + size_of_e = 8 + + # Graph size + G_sz = E * size_of_e + E * size_of_w + V * size_of_v + + # The impact of depth limit depends on the sparsity + # pattern and diameter. We cannot leverage it without + # traversing the full dataset a the moment. + + # dense output + output_sz = n_sources * 2 * V * size_of_v + + # sparse output + if components is not None: + tmp = components["color"].value_counts() + n_components = tmp.size + if n_sources / n_components > 100: + warnings.warn( + "High number of seeds per component result in large output." + ) + mean_component_sz = tmp.mean() + output_sz = mean_component_sz * n_sources * 2 * size_of_e + + # counting 10% for context, handle and temporary allocations + mem_footprint = (G_sz + output_sz) * 1.1 + if mem_footprint > mem: + warnings.warn(f"Cannot execute in-memory :{mem_footprint} Bytes") + + return mem_footprint + + +def concurrent_bfs(Graphs, sources, depth_limit=None, offload=False): + """ + Find the breadth first traversals of multiple graphs with multiple sources + in each graph. + + Parameters + ---------- + Graphs : list of cugraph.Graph or cugraph.DiGraph + The adjacency lists will be computed if not already present. + + sources : list of cudf.Series + For each graph, subset of vertices from which the traversals start. + A BFS is run in Graphs[i] for each source in the Series at sources[i]. + The size of this list must match the size of the graph list. + The size of each Series (ie. the number of sources per graph) + is flexible, but cannot exceed the size of the corresponding graph. + + + depth_limit : Integer, optional, default=None + Limit the depth of the search. Terminates if no more vertices are + reachable within the distance of depth_limit + + offload : boolean, optional, default=False + Indicates if output should be written to the disk. + When not provided, the algorithms decides if offloading is needed + based on the input parameters. + + Returns + ------- + Return type is decided based on the input parameters (size of + sources, size of the graph, number of graphs and offload setting) + + If G is a cugraph.Graph and output fits in memory: + BFS_edge_lists : cudf.DataFrame + GPU data frame containing all BFS edges + source_offsets: cudf.Series + Series containing the starting offset in the returned edge list + for each source. + + If offload is True, or if the output does not fit in memory : + Writes csv files containing BFS output to the disk. + """ + raise NotImplementedError( + "concurrent_bfs is coming soon! Please up vote the github issue 1465\ + to help us prioritize" + ) + if not isinstance(Graphs, list): + raise TypeError( + "Graphs should be a list of cugraph.Graph or cugraph.DiGraph" + ) + if not isinstance(sources, list): + raise TypeError("sources should be a list of cudf.Series") + if len(Graphs) != len(sources): + raise ValueError( + "The size of the sources list must match\ + the size of the graph list." + ) + if offload is True: + raise NotImplementedError( + "Offloading is coming soon! Please up vote the github issue 1461\ + to help us prioritize" + ) + + # Consolidate graphs in a single graph and record components + + # Renumber and concatenate sources in a single df + + # Call multi_source_bfs + # multi_source_bfs( + # G, + # sources, + # components=components, + # depth_limit=depth_limit, + # offload=offload, + # ) + + +def multi_source_bfs( + G, sources, components=None, depth_limit=None, offload=False +): + """ + Find the breadth first traversal from multiple sources in a graph. + + Parameters + ---------- + G : cugraph.Graph or cugraph.DiGraph + The adjacency list will be computed if not already present. + + sources : cudf.Series + Subset of vertices from which the traversals start. A BFS is run for + each source in the Series. + The size of the series should be at least one and cannot exceed the + size of the graph. + + depth_limit : Integer, optional, default=None + Limit the depth of the search. Terminates if no more vertices are + reachable within the distance of depth_limit + + components : cudf.DataFrame, optional, default=None + GPU Dataframe containing the component information. + Passing this information may impact the return type. + When no component information is passed BFS uses one component + behavior settings. + + components['vertex'] : cudf.Series + vertex IDs + components['color'] : cudf.Series + component IDs/color for vertices. + + offload : boolean, optional, default=False + Indicates if output should be written to the disk. + When not provided, the algorithms decides if offloading is needed + based on the input parameters. + + Returns + ------- + Return value type is decided based on the input parameters (size of + sources, size of the graph, number of components and offload setting) + If G is a cugraph.Graph, returns : + cudf.DataFrame + df['vertex'] vertex IDs + + df['distance_'] path distance for each vertex from the + starting vertex. One column per source. + + df['predecessor_'] for each i'th position in the column, + the vertex ID immediately preceding the vertex at position i in + the 'vertex' column. One column per source. + + If G is a cugraph.Graph and component information is present returns : + BFS_edge_lists : cudf.DataFrame + GPU data frame containing all BFS edges + source_offsets: cudf.Series + Series containing the starting offset in the returned edge list + for each source. + + If offload is True, or if the output does not fit in memory : + Writes csv files containing BFS output to the disk. + """ + raise NotImplementedError( + "concurrent_bfs is coming soon! Please up vote the github issue 1465\ + to help us prioritize" + ) + # if components is not None: + # null_check(components["vertex"]) + # null_check(components["colors"]) + # + # if depth_limit is not None: + # raise NotImplementedError( + # "depth limit implementation of BFS is not currently supported" + # ) + + # if offload is True: + # raise NotImplementedError( + # "Offloading is coming soon! Please up vote the github issue 1461 + # to help us prioritize" + # ) + if isinstance(sources, list): + sources = cudf.Series(sources) + if G.renumbered is True: + sources = G.lookup_internal_vertex_id(cudf.Series(sources)) + if not G.adjlist: + G.view_adj_list() + # Memory footprint check + footprint = _get_feasibility( + G, sources, components=components, depth_limit=depth_limit + ) + print(footprint) + # Call multi_source_bfs + # FIXME remove when implemented + # raise NotImplementedError("Commming soon") diff --git a/python/cugraph/utilities/utils.py b/python/cugraph/utilities/utils.py index 39b789d7f79..adaec0f9e44 100644 --- a/python/cugraph/utilities/utils.py +++ b/python/cugraph/utilities/utils.py @@ -26,6 +26,7 @@ from cupyx.scipy.sparse.coo import coo_matrix as cp_coo_matrix from cupyx.scipy.sparse.csr import csr_matrix as cp_csr_matrix from cupyx.scipy.sparse.csc import csc_matrix as cp_csc_matrix + CP_MATRIX_TYPES = [cp_coo_matrix, cp_csr_matrix, cp_csc_matrix] CP_COMPRESSED_MATRIX_TYPES = [cp_csr_matrix, cp_csc_matrix] except ModuleNotFoundError: @@ -38,6 +39,7 @@ from scipy.sparse.coo import coo_matrix as sp_coo_matrix from scipy.sparse.csr import csr_matrix as sp_csr_matrix from scipy.sparse.csc import csc_matrix as sp_csc_matrix + SP_MATRIX_TYPES = [sp_coo_matrix, sp_csr_matrix, sp_csc_matrix] SP_COMPRESSED_MATRIX_TYPES = [sp_csr_matrix, sp_csc_matrix] except ModuleNotFoundError: @@ -80,15 +82,21 @@ def get_traversed_path(df, id): >>> path = cugraph.utils.get_traversed_path(sssp_df, 32) """ - if 'vertex' not in df.columns: - raise ValueError("DataFrame does not appear to be a BFS or " - "SSP result - 'vertex' column missing") - if 'distance' not in df.columns: - raise ValueError("DataFrame does not appear to be a BFS or " - "SSP result - 'distance' column missing") - if 'predecessor' not in df.columns: - raise ValueError("DataFrame does not appear to be a BFS or " - "SSP result - 'predecessor' column missing") + if "vertex" not in df.columns: + raise ValueError( + "DataFrame does not appear to be a BFS or " + "SSP result - 'vertex' column missing" + ) + if "distance" not in df.columns: + raise ValueError( + "DataFrame does not appear to be a BFS or " + "SSP result - 'distance' column missing" + ) + if "predecessor" not in df.columns: + raise ValueError( + "DataFrame does not appear to be a BFS or " + "SSP result - 'predecessor' column missing" + ) if type(id) != int: raise ValueError("The vertex 'id' needs to be an integer") @@ -96,17 +104,17 @@ def get_traversed_path(df, id): # or edited. Therefore we cannot assume that using the vertex ID # as an index will work - ddf = df[df['vertex'] == id] + ddf = df[df["vertex"] == id] if len(ddf) == 0: raise ValueError("The vertex (", id, " is not in the result set") - pred = ddf['predecessor'].iloc[0] + pred = ddf["predecessor"].iloc[0] answer = [] answer.append(ddf) while pred != -1: - ddf = df[df['vertex'] == pred] - pred = ddf['predecessor'].iloc[0] + ddf = df[df["vertex"] == pred] + pred = ddf["predecessor"].iloc[0] answer.append(ddf) return cudf.concat(answer) @@ -138,15 +146,21 @@ def get_traversed_path_list(df, id): >>> path = cugraph.utils.get_traversed_path_list(sssp_df, 32) """ - if 'vertex' not in df.columns: - raise ValueError("DataFrame does not appear to be a BFS or " - "SSP result - 'vertex' column missing") - if 'distance' not in df.columns: - raise ValueError("DataFrame does not appear to be a BFS or " - "SSP result - 'distance' column missing") - if 'predecessor' not in df.columns: - raise ValueError("DataFrame does not appear to be a BFS or " - "SSP result - 'predecessor' column missing") + if "vertex" not in df.columns: + raise ValueError( + "DataFrame does not appear to be a BFS or " + "SSP result - 'vertex' column missing" + ) + if "distance" not in df.columns: + raise ValueError( + "DataFrame does not appear to be a BFS or " + "SSP result - 'distance' column missing" + ) + if "predecessor" not in df.columns: + raise ValueError( + "DataFrame does not appear to be a BFS or " + "SSP result - 'predecessor' column missing" + ) if type(id) != int: raise ValueError("The vertex 'id' needs to be an integer") @@ -158,17 +172,17 @@ def get_traversed_path_list(df, id): answer = [] answer.append(id) - ddf = df[df['vertex'] == id] + ddf = df[df["vertex"] == id] if len(ddf) == 0: raise ValueError("The vertex (", id, " is not in the result set") - pred = ddf['predecessor'].iloc[0] + pred = ddf["predecessor"].iloc[0] while pred != -1: answer.append(pred) - ddf = df[df['vertex'] == pred] - pred = ddf['predecessor'].iloc[0] + ddf = df[df["vertex"] == pred] + pred = ddf["predecessor"].iloc[0] return answer @@ -206,6 +220,14 @@ def is_device_version_less_than(min_version=(7, 0)): return False +def get_device_memory_info(): + """ + Returns the total amount of global memory on the device in bytes + """ + meminfo = cuda.current_context().get_memory_info() + return meminfo[1] + + # FIXME: if G is a Nx type, the weight attribute is assumed to be "weight", if # set. An additional optional parameter for the weight attr name when accepting # Nx graphs may be needed. From the Nx docs: @@ -229,29 +251,35 @@ def ensure_cugraph_obj(obj, nx_weight_attr=None, matrix_graph_type=None): elif (nx is not None) and (input_type in [nx.Graph, nx.DiGraph]): return (convert_from_nx(obj, weight=nx_weight_attr), input_type) - elif (input_type in CP_MATRIX_TYPES) or \ - (input_type in SP_MATRIX_TYPES): + elif (input_type in CP_MATRIX_TYPES) or (input_type in SP_MATRIX_TYPES): if matrix_graph_type is None: matrix_graph_type = Graph elif matrix_graph_type not in [Graph, DiGraph]: - raise TypeError(f"matrix_graph_type must be either a cugraph " - f"Graph or DiGraph, got: {matrix_graph_type}") - - if input_type in (CP_COMPRESSED_MATRIX_TYPES + - SP_COMPRESSED_MATRIX_TYPES): + raise TypeError( + f"matrix_graph_type must be either a cugraph " + f"Graph or DiGraph, got: {matrix_graph_type}" + ) + + if input_type in ( + CP_COMPRESSED_MATRIX_TYPES + SP_COMPRESSED_MATRIX_TYPES + ): coo = obj.tocoo(copy=False) else: coo = obj if input_type in CP_MATRIX_TYPES: - df = cudf.DataFrame({"source": cp.ascontiguousarray(coo.row), - "destination": cp.ascontiguousarray(coo.col), - "weight": cp.ascontiguousarray(coo.data)}) + df = cudf.DataFrame( + { + "source": cp.ascontiguousarray(coo.row), + "destination": cp.ascontiguousarray(coo.col), + "weight": cp.ascontiguousarray(coo.data), + } + ) else: - df = cudf.DataFrame({"source": coo.row, - "destination": coo.col, - "weight": coo.data}) + df = cudf.DataFrame( + {"source": coo.row, "destination": coo.col, "weight": coo.data} + ) # FIXME: # * do a quick check that symmetry is stored explicitly in the cupy # data for sym matrices (ie. for each uv, check vu is there)