void Core::get_dependences_info_clause(PragmaCustomClause clause,
DataSharingEnvironment& data_sharing,
DependencyDirection dep_attr)
{
if (clause.is_defined())
{
ObjectList<Nodecl::NodeclBase> expr_list = clause.get_arguments_as_expressions();
add_data_sharings(expr_list, data_sharing, dep_attr);
}
}
void Core::ompss_common_target_handler_pre(TL::PragmaCustomLine pragma_line,
OmpSs::TargetContext& target_ctx,
TL::Scope scope,
bool is_pragma_task)
{
PragmaCustomClause onto = pragma_line.get_clause("onto");
if (onto.is_defined())
{
target_ctx.onto = onto.get_arguments_as_expressions(scope);
}
PragmaCustomClause device = pragma_line.get_clause("device");
if (device.is_defined())
{
ObjectList<std::string> device_list =
device.get_tokenized_arguments()
.map<const char *>(&std::string::c_str)
.map<std::string>(&strtolower);
target_ctx.device_list.insert(device_list);
}
else
{
// In #pragma omp target a device is mandatory, for #pragma omp task
// add it only if not empty
std::string default_device = "smp";
bool set_default_device = false;
if (!is_pragma_task)
{
warn_printf_at(pragma_line.get_locus(),
"'#pragma omp target' without 'device' clause. Assuming 'device(smp)'\n");
set_default_device = true;
}
else if (target_ctx.device_list.empty())
{
set_default_device = true;
//If onto is defined and there is no device, default device is MPI
if (onto.is_defined()) default_device = "mpi";
}
if (set_default_device)
{
target_ctx.device_list.clear();
target_ctx.device_list.append(default_device);
}
}
PragmaCustomClause copy_in = pragma_line.get_clause("copy_in");
if (copy_in.is_defined())
{
target_ctx.copy_in = parse_dependences_ompss_clause(copy_in, scope);
}
PragmaCustomClause copy_out = pragma_line.get_clause("copy_out");
if (copy_out.is_defined())
{
target_ctx.copy_out = parse_dependences_ompss_clause(copy_out, scope);
}
PragmaCustomClause copy_inout = pragma_line.get_clause("copy_inout");
if (copy_inout.is_defined())
{
target_ctx.copy_inout = parse_dependences_ompss_clause(copy_inout, scope);
}
PragmaCustomClause ndrange = pragma_line.get_clause("ndrange");
if (ndrange.is_defined())
{
target_ctx.ndrange = ndrange.get_arguments_as_expressions(scope);
}
PragmaCustomClause shmem = pragma_line.get_clause("shmem");
if (shmem.is_defined())
{
if (ndrange.is_defined())
{
target_ctx.shmem = shmem.get_arguments_as_expressions(scope);
}
else
{
warn_printf_at(pragma_line.get_locus(), "'shmem' clause cannot be used without the 'ndrange' clause, skipping\n");
}
}
PragmaCustomClause file = pragma_line.get_clause("file");
if (file.is_defined())
{
ObjectList<std::string> file_list = file.get_tokenized_arguments();
if (file_list.size() != 1)
{
warn_printf_at(pragma_line.get_locus(), "clause 'file' expects one identifier, skipping\n");
}
else
{
target_ctx.file = file_list[0];
}
}
PragmaCustomClause name = pragma_line.get_clause("name");
if (name.is_defined())
{
ObjectList<std::string> name_list = name.get_tokenized_arguments();
if (name_list.size() != 1)
{
warn_printf_at(pragma_line.get_locus(), "clause 'name' expects one identifier, skipping\n");
}
else
{
target_ctx.name = name_list[0];
}
}
PragmaCustomClause copy_deps = pragma_line.get_clause("copy_deps");
PragmaCustomClause no_copy_deps = pragma_line.get_clause("no_copy_deps");
if (target_ctx.copy_deps == OmpSs::TargetContext::UNDEF_COPY_DEPS)
{
target_ctx.copy_deps = OmpSs::TargetContext::NO_COPY_DEPS;
if (!copy_deps.is_defined()
&& !no_copy_deps.is_defined())
{
if (this->in_ompss_mode()
&& _copy_deps_by_default)
{
// Copy deps is true only if there is no copy_in, copy_out
// or copy_inout
if ( !copy_in.is_defined()
&& !copy_out.is_defined()
&& !copy_inout.is_defined())
{
target_ctx.copy_deps = OmpSs::TargetContext::COPY_DEPS;
if (!_already_informed_new_ompss_copy_deps)
{
info_printf_at(pragma_line.get_locus(),
"unless 'no_copy_deps' is specified, "
"the default in OmpSs is now 'copy_deps'\n");
info_printf_at(pragma_line.get_locus(),
"this diagnostic is only shown for the "
"first task found\n");
_already_informed_new_ompss_copy_deps = true;
}
}
}
}
else if (copy_deps.is_defined())
{
target_ctx.copy_deps = OmpSs::TargetContext::COPY_DEPS;
}
else if (no_copy_deps.is_defined())
{
target_ctx.copy_deps = OmpSs::TargetContext::NO_COPY_DEPS;
}
else
{
internal_error("Code unreachable", 0);
}
}
else if (target_ctx.copy_deps == OmpSs::TargetContext::NO_COPY_DEPS
|| target_ctx.copy_deps == OmpSs::TargetContext::COPY_DEPS)
{
if (copy_deps.is_defined())
{
warn_printf_at(pragma_line.get_locus(),
"ignoring 'copy_deps' clause because this context is already '%s'\n",
target_ctx.copy_deps == OmpSs::TargetContext::NO_COPY_DEPS ? "no_copy_deps" : "copy_deps");
}
if (no_copy_deps.is_defined())
{
warn_printf_at(pragma_line.get_locus(),
"ignoring 'no_copy_deps' clause because this context is already '%s'\n",
target_ctx.copy_deps == OmpSs::TargetContext::NO_COPY_DEPS ? "no_copy_deps" : "copy_deps");
}
}
else
{
internal_error("Code unreachable", 0);
}
ERROR_CONDITION(target_ctx.copy_deps == OmpSs::TargetContext::UNDEF_COPY_DEPS,
"Invalid value for copy_deps at this point", 0)
PragmaCustomClause implements = pragma_line.get_clause("implements");
if (implements.is_defined())
{
Symbol function_symbol (NULL);
if (IS_C_LANGUAGE
|| IS_CXX_LANGUAGE)
{
ObjectList<Nodecl::NodeclBase> implements_list = implements.get_arguments_as_expressions(scope);
ERROR_CONDITION(implements_list.size() != 1, "clause 'implements' expects one identifier", 0);
Nodecl::NodeclBase implements_name = implements_list[0];
if (implements_name.is<Nodecl::Symbol>())
{
function_symbol = implements_name.get_symbol();
}
else if (implements_name.is<Nodecl::CxxDepNameSimple>())
{
ObjectList<TL::Symbol> symbols = scope.get_symbols_from_name(implements_name.get_text());
ERROR_CONDITION(symbols.size() != 1,
"The argument of the clause 'implements' cannot be an overloaded function identifier", 0);
function_symbol = symbols[0];
}
else
{
internal_error("Unexpected node", 0);
}
}
else if (IS_FORTRAN_LANGUAGE)
{
ObjectList<std::string> implements_list = implements.get_tokenized_arguments();
ERROR_CONDITION(implements_list.size() != 1, "clause 'implements' expects one identifier", 0);
// Restore the scope chain we broke in an INTERFACE block
const decl_context_t* decl_context = scope.get_decl_context();
TL::Symbol current_procedure = scope.get_related_symbol();
decl_context->current_scope->contained_in = current_procedure.get_internal_symbol()->decl_context->current_scope;
TL::Scope fixed_scope = TL::Scope(decl_context);
ObjectList<TL::Symbol> symbols = fixed_scope.get_symbols_from_name(strtolower(implements_list[0].c_str()));
ERROR_CONDITION(symbols.size() != 1,"Unreachable code", 0);
function_symbol = symbols[0];
}
else
{
internal_error("Unreachable code", 0);
}
if (function_symbol.is_valid()
&& function_symbol.is_function())
{
target_ctx.has_implements = true;
target_ctx.implements = function_symbol;
}
else
{
warn_printf_at(pragma_line.get_locus(), "the argument of the clause 'implements' is not a valid identifier, skipping\n");
}
}
}