Skip to content

Commit 74b957d

Browse files
daiyaan-ahmed6daiyaan-ahmed6
authored
[SYCLomatic] Add migration for cudaGraphGetNodes, cudaGraphGetRootNodes (#2600)
Signed-off-by: Daiyaan Ahmed <daiyaan.ahmed@intel.com>
1 parent 950c271 commit 74b957d

6 files changed

Lines changed: 111 additions & 4 deletions

File tree

‎clang/lib/DPCT/RulesLang/APINamesGraph.inc‎

Lines changed: 21 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -76,3 +76,24 @@ CONDITIONAL_FACTORY_ENTRY(
7676
Diagnostics::TRY_EXPERIMENTAL_FEATURE,
7777
ARG("cudaGraphNodeGetType"),
7878
ARG("--use-experimental-features=graph")))
79+
80+
ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
81+
UseExtGraph,
82+
CALL_FACTORY_ENTRY("cudaGraphGetNodes", CALL(MapNames::getDpctNamespace() +
83+
"experimental::get_nodes",
84+
ARG(0), ARG(1), ARG(2))),
85+
UNSUPPORT_FACTORY_ENTRY("cudaGraphGetNodes",
86+
Diagnostics::TRY_EXPERIMENTAL_FEATURE,
87+
ARG("cudaGraphGetNodes"),
88+
ARG("--use-experimental-features=graph"))))
89+
90+
ASSIGNABLE_FACTORY(CONDITIONAL_FACTORY_ENTRY(
91+
UseExtGraph,
92+
CALL_FACTORY_ENTRY("cudaGraphGetRootNodes",
93+
CALL(MapNames::getDpctNamespace() +
94+
"experimental::get_root_nodes",
95+
ARG(0), ARG(1), ARG(2))),
96+
UNSUPPORT_FACTORY_ENTRY("cudaGraphGetRootNodes",
97+
Diagnostics::TRY_EXPERIMENTAL_FEATURE,
98+
ARG("cudaGraphGetRootNodes"),
99+
ARG("--use-experimental-features=graph"))))

‎clang/lib/DPCT/RulesLang/RulesLangGraph.cpp‎

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -33,7 +33,8 @@ void GraphRule::registerMatcher(MatchFinder &MF) {
3333
return hasAnyName("cudaGraphInstantiate", "cudaGraphLaunch",
3434
"cudaGraphExecDestroy", "cudaGraphAddEmptyNode",
3535
"cudaGraphAddDependencies", "cudaGraphExecUpdate",
36-
"cudaGraphNodeGetType");
36+
"cudaGraphNodeGetType", "cudaGraphGetNodes",
37+
"cudaGraphGetRootNodes");
3738
};
3839
MF.addMatcher(
3940
callExpr(callee(functionDecl(functionName()))).bind("FunctionCall"),

‎clang/lib/DPCT/SrcAPI/APINames.inc‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -447,8 +447,8 @@ ENTRY(cudaGraphExternalSemaphoresWaitNodeGetParams, cudaGraphExternalSemaphoresW
447447
ENTRY(cudaGraphExternalSemaphoresWaitNodeSetParams, cudaGraphExternalSemaphoresWaitNodeSetParams, false, NO_FLAG, P4, "comment")
448448
ENTRY(cudaGraphGetEdges, cudaGraphGetEdges, false, NO_FLAG, P4, "comment")
449449
ENTRY(cudaGraphGetEdges_v2, cudaGraphGetEdges_v2, false, NO_FLAG, P4, "comment")
450-
ENTRY(cudaGraphGetNodes, cudaGraphGetNodes, false, NO_FLAG, P4, "comment")
451-
ENTRY(cudaGraphGetRootNodes, cudaGraphGetRootNodes, false, NO_FLAG, P4, "comment")
450+
ENTRY(cudaGraphGetNodes, cudaGraphGetNodes, true, NO_FLAG, P4, "Successful/DPCT1119")
451+
ENTRY(cudaGraphGetRootNodes, cudaGraphGetRootNodes, true, NO_FLAG, P4, "Successful/DPCT1119")
452452
ENTRY(cudaGraphHostNodeGetParams, cudaGraphHostNodeGetParams, false, NO_FLAG, P4, "comment")
453453
ENTRY(cudaGraphHostNodeSetParams, cudaGraphHostNodeSetParams, false, NO_FLAG, P4, "comment")
454454
ENTRY(cudaGraphInstantiate, cudaGraphInstantiate, true, NO_FLAG, P4, "Successful/DPCT1119")

‎clang/runtime/dpct-rt/include/dpct/graph.hpp‎

Lines changed: 58 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -10,6 +10,7 @@
1010

1111
#include <sycl/ext/oneapi/experimental/graph.hpp>
1212
#include <sycl/sycl.hpp>
13+
#include <unordered_map>
1314

1415
namespace dpct {
1516
namespace experimental {
@@ -65,8 +66,42 @@ class graph_mgr {
6566
(*graph)->end_recording();
6667
}
6768

69+
void get_nodes(dpct::experimental::command_graph_ptr graph,
70+
dpct::experimental::node_ptr *nodesArray,
71+
std::size_t *numberOfNodes) {
72+
auto nodes = graph->get_nodes();
73+
nodes_map[graph] = nodes;
74+
*numberOfNodes = nodes.size();
75+
if (!nodesArray) {
76+
return;
77+
}
78+
for (std::size_t i = 0; i < *numberOfNodes; i++) {
79+
nodesArray[i] = &nodes_map[graph][i];
80+
}
81+
}
82+
83+
void get_root_nodes(dpct::experimental::command_graph_ptr graph,
84+
dpct::experimental::node_ptr *nodesArray,
85+
std::size_t *numberOfNodes) {
86+
auto root_nodes = graph->get_root_nodes();
87+
root_nodes_map[graph] = root_nodes;
88+
*numberOfNodes = root_nodes.size();
89+
if (!nodesArray) {
90+
return;
91+
}
92+
for (std::size_t i = 0; i < *numberOfNodes; i++) {
93+
nodesArray[i] = &root_nodes_map[graph][i];
94+
}
95+
}
96+
6897
private:
6998
std::unordered_map<sycl::queue *, command_graph_ptr> queue_graph_map;
99+
std::unordered_map<dpct::experimental::command_graph_ptr,
100+
std::vector<sycl::ext::oneapi::experimental::node>>
101+
nodes_map;
102+
std::unordered_map<dpct::experimental::command_graph_ptr,
103+
std::vector<sycl::ext::oneapi::experimental::node>>
104+
root_nodes_map;
70105
};
71106
} // namespace detail
72107

@@ -133,5 +168,28 @@ static void add_dependencies(dpct::experimental::command_graph_ptr graph,
133168
}
134169
}
135170

171+
/// Gets the nodes in the command graph.
172+
/// \param [in] graph A pointer to the command graph.
173+
/// \param [out] nodesArray An array of node pointers where the
174+
/// nodes will be assigned.
175+
/// \param [out] numberOfNodes The number of nodes in the graph.
176+
static void get_nodes(dpct::experimental::command_graph_ptr graph,
177+
dpct::experimental::node_ptr *nodesArray,
178+
std::size_t *numberOfNodes) {
179+
detail::graph_mgr::instance().get_nodes(graph, nodesArray, numberOfNodes);
180+
}
181+
182+
/// Gets the root nodes in the command graph.
183+
/// \param [in] graph A pointer to the command graph.
184+
/// \param [out] nodesArray An array of node pointers where the
185+
/// root nodes will be assigned.
186+
/// \param [out] numberOfNodes The number of root nodes in the graph.
187+
static void get_root_nodes(dpct::experimental::command_graph_ptr graph,
188+
dpct::experimental::node_ptr *nodesArray,
189+
std::size_t *numberOfNodes) {
190+
detail::graph_mgr::instance().get_root_nodes(graph, nodesArray,
191+
numberOfNodes);
192+
}
193+
136194
} // namespace experimental
137195
} // namespace dpct

‎clang/test/dpct/cudaGraph_test.cu‎

Lines changed: 18 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -70,6 +70,24 @@ int main() {
7070
// CHECK: dpct::experimental::add_empty_node(&node, graph, node10, 1);
7171
cudaGraphAddEmptyNode(&node, graph, node10, 1);
7272

73+
size_t numNodes;
74+
75+
// CHECK: dpct::experimental::get_nodes(graph, node4, &numNodes);
76+
// CHECK-NEXT: CUDA_CHECK_THROW(DPCT_CHECK_ERROR(dpct::experimental::get_nodes(graph, node4, &numNodes)));
77+
cudaGraphGetNodes(graph, node4, &numNodes);
78+
CUDA_CHECK_THROW(cudaGraphGetNodes(graph, node4, &numNodes));
79+
80+
// CHECK: dpct::experimental::get_nodes(*graph2, node5, &numNodes);
81+
cudaGraphGetNodes(*graph2, node5, &numNodes);
82+
83+
// CHECK: dpct::experimental::get_root_nodes(graph, node4, &numNodes);
84+
// CHECK-NEXT: CUDA_CHECK_THROW(DPCT_CHECK_ERROR(dpct::experimental::get_root_nodes(graph, node4, &numNodes)));
85+
cudaGraphGetRootNodes(graph, node4, &numNodes);
86+
CUDA_CHECK_THROW(cudaGraphGetRootNodes(graph, node4, &numNodes));
87+
88+
// CHECK: dpct::experimental::get_root_nodes(*graph2, node5, &numNodes);
89+
cudaGraphGetRootNodes(*graph2, node5, &numNodes);
90+
7391
// CHECK: dpct::experimental::add_dependencies(graph, node4, node5, 10);
7492
// CHECK-NEXT: CUDA_CHECK_THROW(DPCT_CHECK_ERROR(dpct::experimental::add_dependencies(graph, node4, node5, 10)));
7593
cudaGraphAddDependencies(graph, node4, node5, 10);

‎clang/test/dpct/cudaGraph_test_default_option.cu‎

Lines changed: 10 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -51,7 +51,6 @@ int main() {
5151
// CHECK-NEXT: */
5252
captureStatus = cudaStreamCaptureStatusInvalidated;
5353

54-
5554
// CHECK: /*
5655
// CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cudaStreamIsCapturing is not supported, please try to remigrate with option: --use-experimental-features=graph.
5756
// CHECK-NEXT: */
@@ -72,6 +71,16 @@ int main() {
7271
// CHECK-NEXT: */
7372
cudaGraphAddDependencies(graph, NULL, NULL, 0);
7473

74+
// CHECK: /*
75+
// CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cudaGraphGetNodes is not supported, please try to remigrate with option: --use-experimental-features=graph.
76+
// CHECK-NEXT: */
77+
cudaGraphGetNodes(graph, NULL, nullptr);
78+
79+
// CHECK: /*
80+
// CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cudaGraphGetRootNodes is not supported, please try to remigrate with option: --use-experimental-features=graph.
81+
// CHECK-NEXT: */
82+
cudaGraphGetRootNodes(graph, NULL, nullptr);
83+
7584
// CHECK: /*
7685
// CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of cudaGraphInstantiate is not supported, please try to remigrate with option: --use-experimental-features=graph.
7786
// CHECK-NEXT: */

0 commit comments

Comments
 (0)