24408034210

Committed 14 Apr 2026 03:32PM UTC coverage: 64.413% (-0.06%) from 64.469%

Build # 24408034210

Build Type

Pull #679

github

Committed by

web-flow

Commit Message

Merge 7f86463c3 into 505b640ce

Pull Request Pull Request #679: Single GPU BLAS Handle Instantiation

Coverage Stats

9 of 92 new or added lines in 15 files covered. (9.78%)

14 existing lines in 7 files now uncovered.

30559 of 47442 relevant lines covered (64.41%)

583.53 hits per line

Source File
Press 'n' to go to next uncovered line, 'b' for previous

0.0

/opt/src/targets/cuda/stdlib/memset.cpp

#include "sdfg/targets/cuda/stdlib/memset.h"
#include "sdfg/targets/cuda/cuda.h"

namespace sdfg::cuda::stdlib {

MemsetNodeDispatcher_CUDAWithTransfers::MemsetNodeDispatcher_CUDAWithTransfers(
    codegen::LanguageExtension& language_extension,
    const Function& function,
    const data_flow::DataFlowGraph& data_flow_graph,
    const sdfg::stdlib::MemsetNode& node
)
    : codegen::LibraryNodeDispatcher(language_extension, function, data_flow_graph, node) {}

void MemsetNodeDispatcher_CUDAWithTransfers::dispatch_code(
    codegen::PrettyPrinter& stream,
    codegen::PrettyPrinter& globals_stream,
    codegen::CodeSnippetFactory& library_snippet_factory
) {
    auto& node = static_cast<const sdfg::stdlib::MemsetNode&>(node_);

    library_snippet_factory.add_global("#include <cuda.h>");

    stream << "cudaError_t err_cuda;" << std::endl;

    std::string num_expr = language_extension_.expression(node.num());

    stream << "void *d_ptr;" << std::endl;
    stream << "err_cuda = cudaMalloc(&d_ptr, " << num_expr << ");" << std::endl;
    cuda_error_checking(stream, language_extension_, "err_cuda");

    stream << "err_cuda = cudaMemset(d_ptr, " << language_extension_.expression(node.value()) << ", " << num_expr
           << ");" << std::endl;
    cuda_error_checking(stream, language_extension_, "err_cuda");

    stream << "err_cuda = cudaMemcpy(" << node.outputs().at(0) << ", d_ptr, " << num_expr
           << ", cudaMemcpyDeviceToHost);" << std::endl;
    cuda_error_checking(stream, language_extension_, "err_cuda");

    stream << "err_cuda = cudaFree(d_ptr);" << std::endl;
    cuda_error_checking(stream, language_extension_, "err_cuda");
}

MemsetNodeDispatcher_CUDAWithoutTransfers::MemsetNodeDispatcher_CUDAWithoutTransfers(
    codegen::LanguageExtension& language_extension,
    const Function& function,
    const data_flow::DataFlowGraph& data_flow_graph,
    const sdfg::stdlib::MemsetNode& node
)
    : codegen::LibraryNodeDispatcher(language_extension, function, data_flow_graph, node) {}

void MemsetNodeDispatcher_CUDAWithoutTransfers::dispatch_code(
    codegen::PrettyPrinter& stream,
    codegen::PrettyPrinter& globals_stream,
    codegen::CodeSnippetFactory& library_snippet_factory
) {
    auto& node = static_cast<const sdfg::stdlib::MemsetNode&>(node_);

    library_snippet_factory.add_global("#include <cuda.h>");

    stream << "cudaError_t err_cuda;" << std::endl;
    stream << "err_cuda = cudaMemset(" << node.outputs().at(0) << ", " << language_extension_.expression(node.value())
           << ", " << language_extension_.expression(node.num()) << ");" << std::endl;
    cuda_error_checking(stream, language_extension_, "err_cuda");
}

} // namespace sdfg::cuda::stdlib

1	#include "sdfg/targets/cuda/stdlib/memset.h"
2	#include "sdfg/targets/cuda/cuda.h"
3
4	namespace sdfg::cuda::stdlib {
5
6	MemsetNodeDispatcher_CUDAWithTransfers::MemsetNodeDispatcher_CUDAWithTransfers(
7	codegen::LanguageExtension& language_extension,
8	const Function& function,
9	const data_flow::DataFlowGraph& data_flow_graph,
10	const sdfg::stdlib::MemsetNode& node
11	)
12	: codegen::LibraryNodeDispatcher(language_extension, function, data_flow_graph, node) {}	×
13
14	void MemsetNodeDispatcher_CUDAWithTransfers::dispatch_code(
15	codegen::PrettyPrinter& stream,
16	codegen::PrettyPrinter& globals_stream,
17	codegen::CodeSnippetFactory& library_snippet_factory
18	) {	×
19	auto& node = static_cast<const sdfg::stdlib::MemsetNode&>(node_);	×
20
NEW 21	library_snippet_factory.add_global("#include <cuda.h>");	×
22
23	stream << "cudaError_t err_cuda;" << std::endl;	×
24
25	std::string num_expr = language_extension_.expression(node.num());	×
26
27	stream << "void *d_ptr;" << std::endl;	×
28	stream << "err_cuda = cudaMalloc(&d_ptr, " << num_expr << ");" << std::endl;	×
29	cuda_error_checking(stream, language_extension_, "err_cuda");	×
30
31	stream << "err_cuda = cudaMemset(d_ptr, " << language_extension_.expression(node.value()) << ", " << num_expr	×
32	<< ");" << std::endl;	×
33	cuda_error_checking(stream, language_extension_, "err_cuda");	×
34
35	stream << "err_cuda = cudaMemcpy(" << node.outputs().at(0) << ", d_ptr, " << num_expr	×
36	<< ", cudaMemcpyDeviceToHost);" << std::endl;	×
37	cuda_error_checking(stream, language_extension_, "err_cuda");	×
38
39	stream << "err_cuda = cudaFree(d_ptr);" << std::endl;	×
40	cuda_error_checking(stream, language_extension_, "err_cuda");	×
41	}	×
42
43	MemsetNodeDispatcher_CUDAWithoutTransfers::MemsetNodeDispatcher_CUDAWithoutTransfers(
44	codegen::LanguageExtension& language_extension,
45	const Function& function,
46	const data_flow::DataFlowGraph& data_flow_graph,
47	const sdfg::stdlib::MemsetNode& node
48	)
49	: codegen::LibraryNodeDispatcher(language_extension, function, data_flow_graph, node) {}	×
50
51	void MemsetNodeDispatcher_CUDAWithoutTransfers::dispatch_code(
52	codegen::PrettyPrinter& stream,
53	codegen::PrettyPrinter& globals_stream,
54	codegen::CodeSnippetFactory& library_snippet_factory
55	) {	×
56	auto& node = static_cast<const sdfg::stdlib::MemsetNode&>(node_);	×
57
NEW 58	library_snippet_factory.add_global("#include <cuda.h>");	×
59
60	stream << "cudaError_t err_cuda;" << std::endl;	×
61	stream << "err_cuda = cudaMemset(" << node.outputs().at(0) << ", " << language_extension_.expression(node.value())	×
62	<< ", " << language_extension_.expression(node.num()) << ");" << std::endl;	×
63	cuda_error_checking(stream, language_extension_, "err_cuda");	×
64	}	×
65
66	} // namespace sdfg::cuda::stdlib

daisytuner / docc / 24408034210

Source File Press 'n' to go to next uncovered line, 'b' for previous

Source File
Press 'n' to go to next uncovered line, 'b' for previous