AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

SWDEV-439637 - Updated to compile with clang compiler

Browse Source 
Change-Id: Ib0a8e1cc007f083fb1d1f4363cf89ba76ad3c4f2
This commit is contained in:
Anusha GodavarthySurya
2024-01-02 14:56:03 +00:00
committed by Anusha Godavarthy Surya
parent e37aaeae09
commit a1b2cbe44e
6 changed files with 106 additions and 96 deletions
Download Patch File Download Diff File Expand all files Collapse all files
hipamd/src/hip_formatting.hpp
+3 -1
View File
@@ -346,7 +346,9 @@ inline std::ostream& operator<<(std::ostream& os, const hiprtcResult& s) {
} }
inline std::ostream& operator<<(std::ostream& os, const hipJitOption& s) { inline std::ostream& operator<<(std::ostream& os, const hipJitOption& s) {
switch (s) { hiprtcJIT_option option = (hiprtcJIT_option)((int)s);
switch (option) {
case HIPRTC_JIT_MAX_REGISTERS: case HIPRTC_JIT_MAX_REGISTERS:
os << "HIPRTC_JIT_MAX_REGISTERS"; os << "HIPRTC_JIT_MAX_REGISTERS";
break; break;
hipamd/src/hip_graph.cpp
+13 -15
View File
@@ -2295,21 +2295,19 @@ hipError_t hipGraphExecUpdate(hipGraphExec_t hGraphExec, hipGraph_t hGraph,
} }
} }
switch(newGraphNodes[i]->GetType()) { if (newGraphNodes[i]->GetType() == hipGraphNodeTypeMemcpy) {
case hipGraphNodeTypeMemcpy: { // Checks if the memcpy node's parameters are same
// Checks if the memcpy node's parameters are same const hip::GraphMemcpyNode* newMemcpyNode =
const hip::GraphMemcpyNode* newMemcpyNode = static_cast<hip::GraphMemcpyNode const*>(newGraphNodes[i]);
static_cast<hip::GraphMemcpyNode const*>(newGraphNodes[i]); const hip::GraphMemcpyNode* oldMemcpyNode =
const hip::GraphMemcpyNode* oldMemcpyNode = static_cast<hip::GraphMemcpyNode const*>(oldGraphExecNodes[i]);
static_cast<hip::GraphMemcpyNode const*>(oldGraphExecNodes[i]); hipMemcpyKind newKind, oldKind;
hipMemcpyKind newKind, oldKind; newKind = newMemcpyNode->GetMemcpyKind();
newKind = newMemcpyNode->GetMemcpyKind(); oldKind = oldMemcpyNode->GetMemcpyKind();
oldKind = oldMemcpyNode->GetMemcpyKind(); if (newKind != oldKind) {
if (newKind != oldKind) { *hErrorNode_out = reinterpret_cast<hipGraphNode_t>(newGraphNodes[i]);
*hErrorNode_out = reinterpret_cast<hipGraphNode_t>(newGraphNodes[i]); *updateResult_out = hipGraphExecUpdateErrorParametersChanged;
*updateResult_out = hipGraphExecUpdateErrorParametersChanged; HIP_RETURN(hipErrorGraphExecUpdateFailure);
HIP_RETURN(hipErrorGraphExecUpdateFailure);
}
} }
} }
// Checks if all the node's dependencies are same // Checks if all the node's dependencies are same
hipamd/src/hip_graph_internal.hpp
+84 -76
View File
@@ -364,7 +364,9 @@ struct GraphNode : public hipGraphNodeDOTAttribute {
fout << "\"" << fromNodeName << "\" -> \"" << toNodeName << "\"" << std::endl; fout << "\"" << fromNodeName << "\" -> \"" << toNodeName << "\"" << std::endl;
} }
} }
virtual std::string GetLabel(hipGraphDebugDotFlags flag) { return (std::to_string(id_) + "\n" + label_); } virtual std::string GetLabel(hipGraphDebugDotFlags flag) override {
return (std::to_string(id_) + "\n" + label_);
}
unsigned int GetEnabled() const { return isEnabled_; } unsigned int GetEnabled() const { return isEnabled_; }
void SetEnabled(unsigned int isEnabled) { isEnabled_ = isEnabled; } void SetEnabled(unsigned int isEnabled) { isEnabled_ = isEnabled; }
}; };
@@ -669,13 +671,13 @@ struct ChildGraphNode : public GraphNode {
childGraph_ = rhs.childGraph_->clone(); childGraph_ = rhs.childGraph_->clone();
} }
GraphNode* clone() const { GraphNode* clone() const override {
return new ChildGraphNode(static_cast<ChildGraphNode const&>(*this)); return new ChildGraphNode(static_cast<ChildGraphNode const&>(*this));
} }
Graph* GetChildGraph() { return childGraph_; } Graph* GetChildGraph() override { return childGraph_; }
hipError_t GetNumParallelStreams(size_t &num) { hipError_t GetNumParallelStreams(size_t &num) override {
if (false == TopologicalOrder(childGraphNodeOrder_)) { if (false == TopologicalOrder(childGraphNodeOrder_)) {
return hipErrorInvalidValue; return hipErrorInvalidValue;
} }
@@ -690,17 +692,19 @@ struct ChildGraphNode : public GraphNode {
return hipSuccess; return hipSuccess;
} }
void SetStream(hip::Stream* stream, GraphExec* ptr = nullptr) { void SetStream(hip::Stream* stream, GraphExec* ptr = nullptr) override {
stream_ = stream; stream_ = stream;
UpdateStream(parallelLists_, stream, ptr); UpdateStream(parallelLists_, stream, ptr);
} }
// For nodes that are dependent on the child graph node waitlist is the last node of the first // For nodes that are dependent on the child graph node waitlist is the last node of the first
// parallel list // parallel list
std::vector<amd::Command*>& GetCommands() { return parallelLists_[0].back()->GetCommands(); } std::vector<amd::Command*>& GetCommands() override {
return parallelLists_[0].back()->GetCommands();
}
// Create child graph node commands and set waitlists // Create child graph node commands and set waitlists
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -713,20 +717,20 @@ struct ChildGraphNode : public GraphNode {
} }
// //
void UpdateEventWaitLists(amd::Command::EventWaitList waitList) { void UpdateEventWaitLists(amd::Command::EventWaitList waitList) override {
if (startCommand_ != nullptr) { if (startCommand_ != nullptr) {
startCommand_->updateEventWaitList(waitList); startCommand_->updateEventWaitList(waitList);
} }
} }
void GetRunList(std::vector<std::vector<Node>>& parallelList, void GetRunList(std::vector<std::vector<Node>>& parallelList,
std::unordered_map<Node, std::vector<Node>>& dependencies) { std::unordered_map<Node, std::vector<Node>>& dependencies) override {
childGraph_->GetRunList(parallelLists_, nodeWaitLists_); childGraph_->GetRunList(parallelLists_, nodeWaitLists_);
} }
bool TopologicalOrder(std::vector<Node>& TopoOrder) { bool TopologicalOrder(std::vector<Node>& TopoOrder) override {
return childGraph_->TopologicalOrder(TopoOrder); return childGraph_->TopologicalOrder(TopoOrder);
} }
void EnqueueCommands(hipStream_t stream) { void EnqueueCommands(hipStream_t stream) override {
// enqueue child graph start command // enqueue child graph start command
if (startCommand_ != nullptr) { if (startCommand_ != nullptr) {
startCommand_->enqueue(); startCommand_->enqueue();
@@ -755,16 +759,16 @@ struct ChildGraphNode : public GraphNode {
return hipSuccess; return hipSuccess;
} }
hipError_t SetParams(GraphNode* node) { hipError_t SetParams(GraphNode* node) override {
const ChildGraphNode* childGraphNode = static_cast<ChildGraphNode const*>(node); const ChildGraphNode* childGraphNode = static_cast<ChildGraphNode const*>(node);
return SetParams(childGraphNode->childGraph_); return SetParams(childGraphNode->childGraph_);
} }
std::string GetLabel(hipGraphDebugDotFlags flag) { virtual std::string GetLabel(hipGraphDebugDotFlags flag) override {
return std::to_string(GetID()) + "\n" + "graph_" + std::to_string(childGraph_->GetID()); return std::to_string(GetID()) + "\n" + "graph_" + std::to_string(childGraph_->GetID());
} }
virtual void GenerateDOT(std::ostream& fout, hipGraphDebugDotFlags flag) { virtual void GenerateDOT(std::ostream& fout, hipGraphDebugDotFlags flag) override {
childGraph_->GenerateDOT(fout, flag); childGraph_->GenerateDOT(fout, flag);
} }
}; };
@@ -783,7 +787,7 @@ class GraphKernelNode : public GraphNode {
size_t GetKerArgSize() const { return alignedKernArgSize_; } size_t GetKerArgSize() const { return alignedKernArgSize_; }
size_t GetKernargSegmentByteSize() const { return kernargSegmentByteSize_; } size_t GetKernargSegmentByteSize() const { return kernargSegmentByteSize_; }
size_t GetKernargSegmentAlignment() const { return kernargSegmentAlignment_; } size_t GetKernargSegmentAlignment() const { return kernargSegmentAlignment_; }
void PrintAttributes(std::ostream& out, hipGraphDebugDotFlags flag) { void PrintAttributes(std::ostream& out, hipGraphDebugDotFlags flag) override {
out << "["; out << "[";
out << "style"; out << "style";
out << "=\""; out << "=\"";
@@ -817,7 +821,7 @@ class GraphKernelNode : public GraphNode {
} }
} }
std::string GetLabel(hipGraphDebugDotFlags flag) { virtual std::string GetLabel(hipGraphDebugDotFlags flag) override {
hipFunction_t func = getFunc(kernelParams_, ihipGetDevice()); hipFunction_t func = getFunc(kernelParams_, ihipGetDevice());
hip::DeviceFunc* function = hip::DeviceFunc::asFunction(func); hip::DeviceFunc* function = hip::DeviceFunc::asFunction(func);
std::string label; std::string label;
@@ -865,7 +869,7 @@ class GraphKernelNode : public GraphNode {
return label; return label;
} }
std::string GetShape(hipGraphDebugDotFlags flag) { std::string GetShape(hipGraphDebugDotFlags flag) override {
if (flag == hipGraphDebugDotFlagsKernelNodeParams || flag == hipGraphDebugDotFlagsVerbose) { if (flag == hipGraphDebugDotFlagsKernelNodeParams || flag == hipGraphDebugDotFlagsVerbose) {
return "record"; return "record";
} else { } else {
@@ -1003,11 +1007,11 @@ class GraphKernelNode : public GraphNode {
} }
} }
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphKernelNode(static_cast<GraphKernelNode const&>(*this)); return new GraphKernelNode(static_cast<GraphKernelNode const&>(*this));
} }
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
hipFunction_t func = nullptr; hipFunction_t func = nullptr;
hipError_t status = validateKernelParams(&kernelParams_, &func, hipError_t status = validateKernelParams(&kernelParams_, &func,
stream ? hip::getDeviceID(stream->context()) : -1); stream ? hip::getDeviceID(stream->context()) : -1);
@@ -1136,7 +1140,7 @@ class GraphKernelNode : public GraphNode {
return hipSuccess; return hipSuccess;
} }
hipError_t SetParams(GraphNode* node) { hipError_t SetParams(GraphNode* node) override {
const GraphKernelNode* kernelNode = static_cast<GraphKernelNode const*>(node); const GraphKernelNode* kernelNode = static_cast<GraphKernelNode const*>(node);
return SetParams(&kernelNode->kernelParams_); return SetParams(&kernelNode->kernelParams_);
} }
@@ -1184,11 +1188,11 @@ class GraphMemcpyNode : public GraphNode {
copyParams_ = rhs.copyParams_; copyParams_ = rhs.copyParams_;
} }
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphMemcpyNode(static_cast<GraphMemcpyNode const&>(*this)); return new GraphMemcpyNode(static_cast<GraphMemcpyNode const&>(*this));
} }
virtual hipError_t CreateCommand(hip::Stream* stream) { virtual hipError_t CreateCommand(hip::Stream* stream) override {
if ((copyParams_.kind == hipMemcpyHostToHost || copyParams_.kind == hipMemcpyDefault) if ((copyParams_.kind == hipMemcpyHostToHost || copyParams_.kind == hipMemcpyDefault)
&& IsHtoHMemcpy(copyParams_.dstPtr.ptr, copyParams_.srcPtr.ptr)) { && IsHtoHMemcpy(copyParams_.dstPtr.ptr, copyParams_.srcPtr.ptr)) {
return hipSuccess; return hipSuccess;
@@ -1230,14 +1234,14 @@ class GraphMemcpyNode : public GraphNode {
return hipSuccess; return hipSuccess;
} }
virtual hipError_t SetParams(GraphNode* node) { virtual hipError_t SetParams(GraphNode* node) override {
const GraphMemcpyNode* memcpyNode = static_cast<GraphMemcpyNode const*>(node); const GraphMemcpyNode* memcpyNode = static_cast<GraphMemcpyNode const*>(node);
return SetParams(&memcpyNode->copyParams_); return SetParams(&memcpyNode->copyParams_);
} }
// ToDo: use this when commands are cloned and command params are to be updated // ToDo: use this when commands are cloned and command params are to be updated
hipError_t ValidateParams(const hipMemcpy3DParms* pNodeParams); hipError_t ValidateParams(const hipMemcpy3DParms* pNodeParams);
std::string GetLabel(hipGraphDebugDotFlags flag) { virtual std::string GetLabel(hipGraphDebugDotFlags flag) override {
size_t offset = 0; size_t offset = 0;
const HIP_MEMCPY3D pCopy = hip::getDrvMemcpy3DDesc(copyParams_); const HIP_MEMCPY3D pCopy = hip::getDrvMemcpy3DDesc(copyParams_);
hipMemoryType srcMemoryType = pCopy.srcMemoryType; hipMemoryType srcMemoryType = pCopy.srcMemoryType;
@@ -1312,7 +1316,7 @@ class GraphMemcpyNode : public GraphNode {
} }
return label; return label;
} }
std::string GetShape(hipGraphDebugDotFlags flag) { std::string GetShape(hipGraphDebugDotFlags flag) override {
if (flag == hipGraphDebugDotFlagsMemcpyNodeParams || flag == hipGraphDebugDotFlagsVerbose) { if (flag == hipGraphDebugDotFlagsMemcpyNodeParams || flag == hipGraphDebugDotFlagsVerbose) {
return "record"; return "record";
} else { } else {
@@ -1339,11 +1343,11 @@ class GraphMemcpyNode1D : public GraphMemcpyNode {
~GraphMemcpyNode1D() {} ~GraphMemcpyNode1D() {}
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphMemcpyNode1D(static_cast<GraphMemcpyNode1D const&>(*this)); return new GraphMemcpyNode1D(static_cast<GraphMemcpyNode1D const&>(*this));
} }
virtual hipError_t CreateCommand(hip::Stream* stream) { virtual hipError_t CreateCommand(hip::Stream* stream) override {
if ((kind_ == hipMemcpyHostToHost || kind_ == hipMemcpyDefault) && IsHtoHMemcpy(dst_, src_)) { if ((kind_ == hipMemcpyHostToHost || kind_ == hipMemcpyDefault) && IsHtoHMemcpy(dst_, src_)) {
return hipSuccess; return hipSuccess;
} }
@@ -1358,7 +1362,7 @@ class GraphMemcpyNode1D : public GraphMemcpyNode {
return status; return status;
} }
virtual void EnqueueCommands(hipStream_t stream) { virtual void EnqueueCommands(hipStream_t stream) override {
bool isH2H = false; bool isH2H = false;
if ((kind_ == hipMemcpyHostToHost || kind_ == hipMemcpyDefault) && IsHtoHMemcpy(dst_, src_)) { if ((kind_ == hipMemcpyHostToHost || kind_ == hipMemcpyDefault) && IsHtoHMemcpy(dst_, src_)) {
isH2H = true; isH2H = true;
@@ -1419,7 +1423,7 @@ class GraphMemcpyNode1D : public GraphMemcpyNode {
} }
} }
hipMemcpyKind GetMemcpyKind() const { hipMemcpyKind GetMemcpyKind() const override {
return kind_; return kind_;
} }
@@ -1435,13 +1439,13 @@ class GraphMemcpyNode1D : public GraphMemcpyNode {
return hipSuccess; return hipSuccess;
} }
virtual hipError_t SetParams(GraphNode* node) { virtual hipError_t SetParams(GraphNode* node) override {
const GraphMemcpyNode1D* memcpy1DNode = static_cast<GraphMemcpyNode1D const*>(node); const GraphMemcpyNode1D* memcpy1DNode = static_cast<GraphMemcpyNode1D const*>(node);
return SetParams(memcpy1DNode->dst_, memcpy1DNode->src_, memcpy1DNode->count_, return SetParams(memcpy1DNode->dst_, memcpy1DNode->src_, memcpy1DNode->count_,
memcpy1DNode->kind_); memcpy1DNode->kind_);
} }
static hipError_t ValidateParams(void* dst, const void* src, size_t count, hipMemcpyKind kind); static hipError_t ValidateParams(void* dst, const void* src, size_t count, hipMemcpyKind kind);
std::string GetLabel(hipGraphDebugDotFlags flag) { virtual std::string GetLabel(hipGraphDebugDotFlags flag) override {
size_t sOffsetOrig = 0; size_t sOffsetOrig = 0;
amd::Memory* origSrcMemory = getMemoryObject(src_, sOffsetOrig); amd::Memory* origSrcMemory = getMemoryObject(src_, sOffsetOrig);
size_t dOffsetOrig = 0; size_t dOffsetOrig = 0;
@@ -1468,16 +1472,18 @@ class GraphMemcpyNode1D : public GraphMemcpyNode {
std::string label; std::string label;
if (flag == hipGraphDebugDotFlagsMemcpyNodeParams || flag == hipGraphDebugDotFlagsVerbose) { if (flag == hipGraphDebugDotFlagsMemcpyNodeParams || flag == hipGraphDebugDotFlagsVerbose) {
char buffer[500]; char buffer[500];
sprintf(buffer, sprintf(
"{\n%s\n| {{ID | node handle} | {%u | %p}}\n| {kind | %s}\n| {{srcPtr | dstPtr} | " buffer,
"{pitch " "{\n%s\n| {{ID | node handle} | {%u | %p}}\n| {kind | %s}\n| {{srcPtr | dstPtr} | "
"| ptr | xsize | ysize | pitch | ptr | xsize | size} | {%zu | %p | %zu | %zu | %zu | %p " "{pitch "
"| %zu " "| ptr | xsize | ysize | pitch | ptr | xsize | size} | {%zu | %p | %zu | %zu | %zu | %p "
"| %zu}}\n| {{srcPos | {{x | %zu} | {y | %zu} | {z | %zu}}} | {dstPos | {{x | %zu} | {y " "| %zu "
"| " "| %zu}}\n| {{srcPos | {{x | %zu} | {y | %zu} | {z | %zu}}} | {dstPos | {{x | %zu} | {y "
"%zu} | {z | %zu}}} | {Extent | {{Width | %zu} | {Height | %zu} | {Depth | %zu}}}}\n}", "| "
label_.c_str(), GetID(), this, memcpyDirection.c_str(), (size_t)0, "%zu} | {z | %zu}}} | {Extent | {{Width | %zu} | {Height | %zu} | {Depth | %zu}}}}\n}",
src_, (size_t)0, (size_t)0, (size_t)0, dst_, (size_t)0, (size_t)0, (size_t)0, (size_t)0, (size_t)0, (size_t)0, (size_t)0, (size_t)0, count_, (size_t)1, (size_t)1); label_.c_str(), GetID(), this, memcpyDirection.c_str(), (size_t)0, src_, (size_t)0,
(size_t)0, (size_t)0, dst_, (size_t)0, (size_t)0, (size_t)0, (size_t)0, (size_t)0,
(size_t)0, (size_t)0, (size_t)0, count_, (size_t)1, (size_t)1);
label = buffer; label = buffer;
} else { } else {
label = std::to_string(GetID()) + "\n" + label_ + "\n(" + memcpyDirection + "," + label = std::to_string(GetID()) + "\n" + label_ + "\n(" + memcpyDirection + "," +
@@ -1485,7 +1491,7 @@ class GraphMemcpyNode1D : public GraphMemcpyNode {
} }
return label; return label;
} }
std::string GetShape(hipGraphDebugDotFlags flag) { std::string GetShape(hipGraphDebugDotFlags flag) override {
if (flag == hipGraphDebugDotFlagsMemcpyNodeParams || flag == hipGraphDebugDotFlagsVerbose) { if (flag == hipGraphDebugDotFlagsMemcpyNodeParams || flag == hipGraphDebugDotFlagsVerbose) {
return "record"; return "record";
} else { } else {
@@ -1507,12 +1513,12 @@ class GraphMemcpyNodeFromSymbol : public GraphMemcpyNode1D {
~GraphMemcpyNodeFromSymbol() {} ~GraphMemcpyNodeFromSymbol() {}
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphMemcpyNodeFromSymbol( return new GraphMemcpyNodeFromSymbol(
static_cast<GraphMemcpyNodeFromSymbol const&>(*this)); static_cast<GraphMemcpyNodeFromSymbol const&>(*this));
} }
virtual hipError_t CreateCommand(hip::Stream* stream) { virtual hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -1580,7 +1586,7 @@ class GraphMemcpyNodeFromSymbol : public GraphMemcpyNode1D {
return hipSuccess; return hipSuccess;
} }
virtual hipError_t SetParams(GraphNode* node) { virtual hipError_t SetParams(GraphNode* node) override {
const GraphMemcpyNodeFromSymbol* memcpyNode = const GraphMemcpyNodeFromSymbol* memcpyNode =
static_cast<GraphMemcpyNodeFromSymbol const*>(node); static_cast<GraphMemcpyNodeFromSymbol const*>(node);
return SetParams(memcpyNode->dst_, memcpyNode->symbol_, memcpyNode->count_, memcpyNode->offset_, return SetParams(memcpyNode->dst_, memcpyNode->symbol_, memcpyNode->count_, memcpyNode->offset_,
@@ -1600,11 +1606,11 @@ class GraphMemcpyNodeToSymbol : public GraphMemcpyNode1D {
~GraphMemcpyNodeToSymbol() {} ~GraphMemcpyNodeToSymbol() {}
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphMemcpyNodeToSymbol(static_cast<GraphMemcpyNodeToSymbol const&>(*this)); return new GraphMemcpyNodeToSymbol(static_cast<GraphMemcpyNodeToSymbol const&>(*this));
} }
virtual hipError_t CreateCommand(hip::Stream* stream) { virtual hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -1670,7 +1676,7 @@ class GraphMemcpyNodeToSymbol : public GraphMemcpyNode1D {
return hipSuccess; return hipSuccess;
} }
virtual hipError_t SetParams(GraphNode* node) { virtual hipError_t SetParams(GraphNode* node) override {
const GraphMemcpyNodeToSymbol* memcpyNode = const GraphMemcpyNodeToSymbol* memcpyNode =
static_cast<GraphMemcpyNodeToSymbol const*>(node); static_cast<GraphMemcpyNodeToSymbol const*>(node);
return SetParams(memcpyNode->src_, memcpyNode->symbol_, memcpyNode->count_, memcpyNode->offset_, return SetParams(memcpyNode->src_, memcpyNode->symbol_, memcpyNode->count_, memcpyNode->offset_,
@@ -1698,11 +1704,11 @@ class GraphMemsetNode : public GraphNode {
memsetParams_ = memsetNode.memsetParams_; memsetParams_ = memsetNode.memsetParams_;
} }
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphMemsetNode(static_cast<GraphMemsetNode const&>(*this)); return new GraphMemsetNode(static_cast<GraphMemsetNode const&>(*this));
} }
std::string GetLabel(hipGraphDebugDotFlags flag) { virtual std::string GetLabel(hipGraphDebugDotFlags flag) override {
std::string label; std::string label;
if (flag == hipGraphDebugDotFlagsMemsetNodeParams || flag == hipGraphDebugDotFlagsVerbose) { if (flag == hipGraphDebugDotFlagsMemsetNodeParams || flag == hipGraphDebugDotFlagsVerbose) {
char buffer[500]; char buffer[500];
@@ -1726,7 +1732,7 @@ class GraphMemsetNode : public GraphNode {
return label; return label;
} }
std::string GetShape(hipGraphDebugDotFlags flag) { std::string GetShape(hipGraphDebugDotFlags flag) override {
if (flag == hipGraphDebugDotFlagsMemsetNodeParams || flag == hipGraphDebugDotFlagsVerbose) { if (flag == hipGraphDebugDotFlagsMemsetNodeParams || flag == hipGraphDebugDotFlagsVerbose) {
return "record"; return "record";
} else { } else {
@@ -1734,7 +1740,7 @@ class GraphMemsetNode : public GraphNode {
} }
} }
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -1748,7 +1754,9 @@ class GraphMemsetNode : public GraphNode {
commands_, commands_,
{memsetParams_.dst, memsetParams_.pitch, memsetParams_.width * memsetParams_.elementSize, {memsetParams_.dst, memsetParams_.pitch, memsetParams_.width * memsetParams_.elementSize,
memsetParams_.height}, memsetParams_.height},
memsetParams_.value, {memsetParams_.width * memsetParams_.elementSize, memsetParams_.height, 1}, stream, memsetParams_.elementSize); memsetParams_.value,
{memsetParams_.width * memsetParams_.elementSize, memsetParams_.height, 1}, stream,
memsetParams_.elementSize);
} }
return status; return status;
} }
@@ -1818,9 +1826,9 @@ class GraphMemsetNode : public GraphNode {
} }
} }
sizeBytes = params->width * params->elementSize * params->height * 1; sizeBytes = params->width * params->elementSize * params->height * 1;
hip_error = hip_error = ihipMemset3D_validate(
ihipMemset3D_validate({params->dst, params->pitch, params->width * params->elementSize, params->height}, {params->dst, params->pitch, params->width * params->elementSize, params->height},
params->value, {params->width * params->elementSize, params->height, 1}, sizeBytes); params->value, {params->width * params->elementSize, params->height, 1}, sizeBytes);
} }
if (hip_error != hipSuccess) { if (hip_error != hipSuccess) {
return hip_error; return hip_error;
@@ -1841,7 +1849,7 @@ class GraphMemsetNode : public GraphNode {
pmemsetParams.width = params->width; pmemsetParams.width = params->width;
return SetParamsInternal(&pmemsetParams, isExec); return SetParamsInternal(&pmemsetParams, isExec);
} }
hipError_t SetParams(GraphNode* node) { hipError_t SetParams(GraphNode* node) override {
const GraphMemsetNode* memsetNode = static_cast<GraphMemsetNode const*>(node); const GraphMemsetNode* memsetNode = static_cast<GraphMemsetNode const*>(node);
return SetParams(&memsetNode->memsetParams_); return SetParams(&memsetNode->memsetParams_);
} }
@@ -1856,11 +1864,11 @@ class GraphEventRecordNode : public GraphNode {
event_(event) {} event_(event) {}
~GraphEventRecordNode() {} ~GraphEventRecordNode() {}
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphEventRecordNode(static_cast<GraphEventRecordNode const&>(*this)); return new GraphEventRecordNode(static_cast<GraphEventRecordNode const&>(*this));
} }
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -1873,7 +1881,7 @@ class GraphEventRecordNode : public GraphNode {
return status; return status;
} }
void EnqueueCommands(hipStream_t stream) { void EnqueueCommands(hipStream_t stream) override {
if (!commands_.empty()) { if (!commands_.empty()) {
hip::Event* e = reinterpret_cast<hip::Event*>(event_); hip::Event* e = reinterpret_cast<hip::Event*>(event_);
// command release during enqueueRecordCommand // command release during enqueueRecordCommand
@@ -1893,7 +1901,7 @@ class GraphEventRecordNode : public GraphNode {
return hipSuccess; return hipSuccess;
} }
hipError_t SetParams(GraphNode* node) { hipError_t SetParams(GraphNode* node) override {
const GraphEventRecordNode* eventRecordNode = const GraphEventRecordNode* eventRecordNode =
static_cast<GraphEventRecordNode const*>(node); static_cast<GraphEventRecordNode const*>(node);
return SetParams(eventRecordNode->event_); return SetParams(eventRecordNode->event_);
@@ -1909,11 +1917,11 @@ class GraphEventWaitNode : public GraphNode {
event_(event) {} event_(event) {}
~GraphEventWaitNode() {} ~GraphEventWaitNode() {}
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphEventWaitNode(static_cast<GraphEventWaitNode const&>(*this)); return new GraphEventWaitNode(static_cast<GraphEventWaitNode const&>(*this));
} }
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -1926,7 +1934,7 @@ class GraphEventWaitNode : public GraphNode {
return status; return status;
} }
void EnqueueCommands(hipStream_t stream) { void EnqueueCommands(hipStream_t stream) override {
if (!commands_.empty()) { if (!commands_.empty()) {
hip::Event* e = reinterpret_cast<hip::Event*>(event_); hip::Event* e = reinterpret_cast<hip::Event*>(event_);
hipError_t status = e->enqueueStreamWaitCommand(stream, commands_[0]); hipError_t status = e->enqueueStreamWaitCommand(stream, commands_[0]);
@@ -1946,7 +1954,7 @@ class GraphEventWaitNode : public GraphNode {
return hipSuccess; return hipSuccess;
} }
hipError_t SetParams(GraphNode* node) { hipError_t SetParams(GraphNode* node) override {
const GraphEventWaitNode* eventWaitNode = static_cast<GraphEventWaitNode const*>(node); const GraphEventWaitNode* eventWaitNode = static_cast<GraphEventWaitNode const*>(node);
return SetParams(eventWaitNode->event_); return SetParams(eventWaitNode->event_);
} }
@@ -1966,11 +1974,11 @@ class GraphHostNode : public GraphNode {
NodeParams_ = hostNode.NodeParams_; NodeParams_ = hostNode.NodeParams_;
} }
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphHostNode(static_cast<GraphHostNode const&>(*this)); return new GraphHostNode(static_cast<GraphHostNode const&>(*this));
} }
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -1987,7 +1995,7 @@ class GraphHostNode : public GraphNode {
NodeParams->fn(NodeParams->userData); NodeParams->fn(NodeParams->userData);
} }
void EnqueueCommands(hipStream_t stream) { void EnqueueCommands(hipStream_t stream) override {
if (!commands_.empty()) { if (!commands_.empty()) {
if (!commands_[0]->setCallback(CL_COMPLETE, GraphHostNode::Callback, &NodeParams_)) { if (!commands_[0]->setCallback(CL_COMPLETE, GraphHostNode::Callback, &NodeParams_)) {
ClPrint(amd::LOG_ERROR, amd::LOG_CODE, "[hipGraph] Failed during setCallback"); ClPrint(amd::LOG_ERROR, amd::LOG_CODE, "[hipGraph] Failed during setCallback");
@@ -2015,7 +2023,7 @@ class GraphHostNode : public GraphNode {
return hipSuccess; return hipSuccess;
} }
hipError_t SetParams(GraphNode* node) { hipError_t SetParams(GraphNode* node) override {
const GraphHostNode* hostNode = static_cast<GraphHostNode const*>(node); const GraphHostNode* hostNode = static_cast<GraphHostNode const*>(node);
return SetParams(&hostNode->NodeParams_); return SetParams(&hostNode->NodeParams_);
} }
@@ -2026,11 +2034,11 @@ class GraphEmptyNode : public GraphNode {
GraphEmptyNode() : GraphNode(hipGraphNodeTypeEmpty, "solid", "rectangle", "EMPTY") {} GraphEmptyNode() : GraphNode(hipGraphNodeTypeEmpty, "solid", "rectangle", "EMPTY") {}
~GraphEmptyNode() {} ~GraphEmptyNode() {}
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphEmptyNode(static_cast<GraphEmptyNode const&>(*this)); return new GraphEmptyNode(static_cast<GraphEmptyNode const&>(*this));
} }
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
hipError_t status = GraphNode::CreateCommand(stream); hipError_t status = GraphNode::CreateCommand(stream);
if (status != hipSuccess) { if (status != hipSuccess) {
return status; return status;
@@ -2044,8 +2052,8 @@ class GraphEmptyNode : public GraphNode {
}; };
// ================================================================================================ // ================================================================================================
class GraphMemAllocNode : public GraphNode { class GraphMemAllocNode final : public GraphNode {
hipMemAllocNodeParams node_params_; // Node parameters for memory allocation hipMemAllocNodeParams node_params_; // Node parameters for memory allocation
amd::Memory* va_ = nullptr; // Memory object, which holds a virtual address amd::Memory* va_ = nullptr; // Memory object, which holds a virtual address
// Derive the new class for VirtualMapCommand, // Derive the new class for VirtualMapCommand,
@@ -2289,11 +2297,11 @@ class GraphDrvMemcpyNode : public GraphNode {
copyParams_ = rhs.copyParams_; copyParams_ = rhs.copyParams_;
} }
GraphNode* clone() const { GraphNode* clone() const override {
return new GraphDrvMemcpyNode(static_cast<GraphDrvMemcpyNode const&>(*this)); return new GraphDrvMemcpyNode(static_cast<GraphDrvMemcpyNode const&>(*this));
} }
hipError_t CreateCommand(hip::Stream* stream) { hipError_t CreateCommand(hip::Stream* stream) override {
if(copyParams_.srcMemoryType == hipMemoryTypeHost && if(copyParams_.srcMemoryType == hipMemoryTypeHost &&
copyParams_.dstMemoryType == hipMemoryTypeHost && copyParams_.dstMemoryType == hipMemoryTypeHost &&
IsHtoHMemcpy(copyParams_.dstHost, copyParams_.srcHost)) { IsHtoHMemcpy(copyParams_.dstHost, copyParams_.srcHost)) {
@@ -2337,7 +2345,7 @@ class GraphDrvMemcpyNode : public GraphNode {
std::memcpy(&copyParams_, params, sizeof(HIP_MEMCPY3D)); std::memcpy(&copyParams_, params, sizeof(HIP_MEMCPY3D));
return hipSuccess; return hipSuccess;
} }
hipError_t SetParams(GraphNode* node) { hipError_t SetParams(GraphNode* node) override {
const GraphDrvMemcpyNode* memcpyNode = static_cast<GraphDrvMemcpyNode const*>(node); const GraphDrvMemcpyNode* memcpyNode = static_cast<GraphDrvMemcpyNode const*>(node);
return SetParams(&memcpyNode->copyParams_); return SetParams(&memcpyNode->copyParams_);
} }
hipamd/src/hip_hmm.cpp
+1 -1
View File
@@ -232,7 +232,7 @@ hipError_t hipStreamAttachMemAsync(hipStream_t stream, void* dev_ptr,
size_t length, unsigned int flags) { size_t length, unsigned int flags) {
HIP_INIT_API(hipStreamAttachMemAsync, stream, dev_ptr, length, flags); HIP_INIT_API(hipStreamAttachMemAsync, stream, dev_ptr, length, flags);
// stream can be null, length can be 0. // stream can be null, length can be 0.
if ((dev_ptr == nullptr)) { if (dev_ptr == nullptr) {
HIP_RETURN(hipErrorInvalidValue); HIP_RETURN(hipErrorInvalidValue);
} }
hipamd/src/hip_module.cpp
+1 -1
View File
@@ -115,7 +115,7 @@ hipError_t hipModuleGetGlobal(hipDeviceptr_t* dptr, size_t* bytes, hipModule_t h
hipError_t hipFuncGetAttribute(int* value, hipFunction_attribute attrib, hipFunction_t hfunc) { hipError_t hipFuncGetAttribute(int* value, hipFunction_attribute attrib, hipFunction_t hfunc) {
HIP_INIT_API(hipFuncGetAttribute, value, attrib, hfunc); HIP_INIT_API(hipFuncGetAttribute, value, attrib, hfunc);
if ((value == nullptr)) { if (value == nullptr) {
HIP_RETURN(hipErrorInvalidValue); HIP_RETURN(hipErrorInvalidValue);
} }
hipamd/src/hip_table_interface.cpp
+4 -2
View File
@@ -131,10 +131,12 @@ hipError_t hipCreateTextureObject(hipTextureObject_t* pTexObject, const hipResou
return hip::GetHipDispatchTable()->hipCreateTextureObject_fn(pTexObject, pResDesc, pTexDesc, return hip::GetHipDispatchTable()->hipCreateTextureObject_fn(pTexObject, pResDesc, pTexDesc,
pResViewDesc); pResViewDesc);
} }
hipError_t hipCtxCreate(hipCtx_t* ctx, unsigned int flags, hipDevice_t device) { extern "C" hipError_t hipCtxCreate(hipCtx_t* ctx, unsigned int flags, hipDevice_t device) {
return hip::GetHipDispatchTable()->hipCtxCreate_fn(ctx, flags, device); return hip::GetHipDispatchTable()->hipCtxCreate_fn(ctx, flags, device);
} }
hipError_t hipCtxDestroy(hipCtx_t ctx) { return hip::GetHipDispatchTable()->hipCtxDestroy_fn(ctx); } extern "C" hipError_t hipCtxDestroy(hipCtx_t ctx) {
return hip::GetHipDispatchTable()->hipCtxDestroy_fn(ctx);
}
hipError_t hipCtxDisablePeerAccess(hipCtx_t peerCtx) { hipError_t hipCtxDisablePeerAccess(hipCtx_t peerCtx) {
return hip::GetHipDispatchTable()->hipCtxDisablePeerAccess_fn(peerCtx); return hip::GetHipDispatchTable()->hipCtxDisablePeerAccess_fn(peerCtx);
} }