示例#1
0
    void Core::get_dependences_info(TL::PragmaCustomLine construct, DataSharingEnvironment& data_sharing)
    {
        PragmaCustomClause input_clause = construct.get_clause("in",
                /* deprecated */ "input");
        get_dependences_info_clause(input_clause, data_sharing, DEP_DIR_IN);

        PragmaCustomClause output_clause = construct.get_clause("out",
                /* deprecated */ "output");
        get_dependences_info_clause(output_clause, data_sharing, DEP_DIR_OUT);

        PragmaCustomClause inout_clause = construct.get_clause("inout");
        get_dependences_info_clause(inout_clause, data_sharing, DEP_DIR_INOUT);

        PragmaCustomClause concurrent_clause = construct.get_clause("concurrent");
        get_dependences_info_clause(concurrent_clause, data_sharing,
                DEP_CONCURRENT);

        PragmaCustomClause commutative_clause = construct.get_clause("commutative");
        get_dependences_info_clause(commutative_clause, data_sharing,
                DEP_COMMUTATIVE);

        // OpenMP standard proposal
        PragmaCustomClause depends = construct.get_clause("depend");
        get_dependences_info_std_clause(construct, depends, data_sharing);
    }
示例#2
0
        // This function is invoked only for inline tasks (and some other
        // constructs though target info is unused for them)
        void Core::ompss_get_target_info(TL::PragmaCustomLine construct,
                DataEnvironment& data_sharing_environment)
        {
            if (_target_context.empty())
                return;

            TL::OmpSs::TargetInfo target_info;

            TL::Symbol enclosing_function = Nodecl::Utils::get_enclosing_function(construct);
            ERROR_CONDITION(!enclosing_function.is_valid(), "This symbol is not valid", 0);
            target_info.set_target_symbol(enclosing_function);
            OmpSs::TargetContext& target_ctx = _target_context.top();

            add_copy_items(construct, data_sharing_environment,
                    target_ctx.copy_in,
                    TL::OmpSs::COPY_DIR_IN,
                    target_info,
                    in_ompss_mode());

            add_copy_items(construct, data_sharing_environment,
                    target_ctx.copy_out,
                    TL::OmpSs::COPY_DIR_OUT,
                    target_info,
                    in_ompss_mode());

            add_copy_items(construct, data_sharing_environment,
                    target_ctx.copy_inout,
                    TL::OmpSs::COPY_DIR_INOUT,
                    target_info,
                    in_ompss_mode());

            target_info.set_file(target_ctx.file);
            target_info.set_name(target_ctx.name);
            target_info.append_to_ndrange(target_ctx.ndrange);
            target_info.append_to_shmem(target_ctx.shmem);
            target_info.append_to_onto(target_ctx.onto);
            target_info.append_to_device_list(target_ctx.device_list);

            // Set data sharings for referenced entities in copies
            if (target_ctx.copy_deps == OmpSs::TargetContext::COPY_DEPS)
            {
                // Copy the dependences, as well

                ObjectList<DependencyItem> dependences;
                data_sharing_environment.get_all_dependences(dependences);

                ObjectList<Nodecl::NodeclBase> dep_list_in;
                ObjectList<Nodecl::NodeclBase> dep_list_out;
                ObjectList<Nodecl::NodeclBase> dep_list_inout;
                ObjectList<Nodecl::NodeclBase> dep_list_weakin;
                ObjectList<Nodecl::NodeclBase> dep_list_weakout;
                ObjectList<Nodecl::NodeclBase> dep_list_weakinout;
                ObjectList<Nodecl::NodeclBase> dep_list_reductions;
                for (ObjectList<DependencyItem>::iterator it = dependences.begin();
                        it != dependences.end();
                        it++)
                {
                    ObjectList<Nodecl::NodeclBase>* p = NULL;
                    switch (it->get_kind())
                    {
                        case DEP_DIR_IN:
                        case DEP_OMPSS_DIR_IN_PRIVATE:
                            {
                                p = &dep_list_in;
                                break;
                            }
                        case DEP_DIR_OUT:
                            {
                                p = &dep_list_out;
                                break;
                            }
                        case DEP_DIR_INOUT:
                            // OmpSs
                        case DEP_OMPSS_CONCURRENT:
                        case DEP_OMPSS_COMMUTATIVE:
                        case DEP_OMPSS_REDUCTION:
                            {
                                p = &dep_list_inout;
                                break;
                            }
                        case DEP_OMPSS_WEAK_IN:
                            {
                                p = &dep_list_weakin;
                                break;
                            }
                        case DEP_OMPSS_WEAK_OUT:
                            {
                                p = &dep_list_weakout;
                                break;
                            }
                        case DEP_OMPSS_WEAK_INOUT:
                        case DEP_OMPSS_WEAK_COMMUTATIVE:
                            {
                                p = &dep_list_weakinout;
                                break;
                            }
                        default:
                            {
                                internal_error("Invalid dependency kind", 0);
                            }
                    }

                    p->append(*it);
                }

                add_copy_items(construct, data_sharing_environment,
                        dep_list_in,
                        TL::OmpSs::COPY_DIR_IN,
                        target_info,
                        in_ompss_mode());

                add_copy_items(construct, data_sharing_environment,
                        dep_list_out,
                        TL::OmpSs::COPY_DIR_OUT,
                        target_info,
                        in_ompss_mode());

                add_copy_items(construct, data_sharing_environment,
                        dep_list_inout,
                        TL::OmpSs::COPY_DIR_INOUT,
                        target_info,
                        in_ompss_mode());

                if (!dep_list_weakin.empty()
                        || !dep_list_weakout.empty()
                        || !dep_list_weakinout.empty())
                {
                    warn_printf_at(construct.get_locus(),
                            "weak dependences are not considered yet for copy_deps\n");
                }
            }

            if (this->in_ompss_mode()
                    && (target_ctx.copy_deps == OmpSs::TargetContext::COPY_DEPS
                        || !target_ctx.copy_in.empty()
                        || !target_ctx.copy_out.empty()
                        || !target_ctx.copy_inout.empty())
                    && !_allow_shared_without_copies)
            {
                ObjectList<TL::OmpSs::CopyItem> all_copies;
                all_copies.append(target_info.get_copy_in());
                all_copies.append(target_info.get_copy_out());
                all_copies.append(target_info.get_copy_inout());

                ObjectList<Symbol> all_copied_syms = all_copies.map<TL::Symbol>(&DataReference::get_base_symbol);

                // In devices with disjoint memory, it may be wrong to use a
                // global variables inside a pragma task without copying it.
                ObjectList<Symbol> ds_syms;
                data_sharing_environment.get_all_symbols(DS_SHARED, ds_syms);

                for(ObjectList<Symbol>::iterator io_it = ds_syms.begin(); 
                        io_it != ds_syms.end(); 
                        io_it++)
                {
                    // Ignore 'this'
                    if (IS_CXX_LANGUAGE
                            && io_it->get_name() == "this")
                    {
                        continue;
                    }

                    if (!all_copied_syms.contains(*io_it))
                    {
                        warn_printf_at(
                                construct.get_locus(),
                                "symbol '%s' has shared data-sharing but does not have"
                                " copy directionality. This may cause problems at run-time\n",
                                io_it->get_qualified_name().c_str());
                    }
                }
            }

            // Store the target information in the current data sharing
            data_sharing_environment.set_target_info(target_info);
        }
示例#3
0
        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");
                }
            }
        }