Пример #1
0
void FindActionParameters::postorder(const IR::ActionListElement* element) {
    auto path = element->getPath();
    auto decl = refMap->getDeclaration(path, true);
    BUG_CHECK(decl->is<IR::P4Action>(), "%1%: not an action", element);
    BUG_CHECK(element->expression->is<IR::MethodCallExpression>(),
              "%1%: expected a method call", element->expression);
    invocations->bind(decl->to<IR::P4Action>(),
                      element->expression->to<IR::MethodCallExpression>(), false);
}
Пример #2
0
const IR::Node* DoRemoveActionParameters::postorder(IR::P4Action* action) {
    LOG1("Visiting " << dbp(action));
    BUG_CHECK(getParent<IR::P4Control>() || getParent<IR::P4Program>(),
              "%1%: unexpected parent %2%", getOriginal(), getContext()->node);
    auto result = new IR::IndexedVector<IR::Declaration>();
    auto leftParams = new IR::IndexedVector<IR::Parameter>();
    auto initializers = new IR::IndexedVector<IR::StatOrDecl>();
    auto postamble = new IR::IndexedVector<IR::StatOrDecl>();
    auto invocation = invocations->get(getOriginal<IR::P4Action>());
    if (invocation == nullptr)
        return action;
    auto args = invocation->arguments;

    ParameterSubstitution substitution;
    substitution.populate(action->parameters, args);

    bool removeAll = invocations->removeAllParameters(getOriginal<IR::P4Action>());
    for (auto p : action->parameters->parameters) {
        if (p->direction == IR::Direction::None && !removeAll) {
            leftParams->push_back(p);
        } else {
            auto decl = new IR::Declaration_Variable(p->srcInfo, p->name,
                                                     p->annotations, p->type, nullptr);
            LOG3("Added declaration " << decl << " annotations " << p->annotations);
            result->push_back(decl);
            auto arg = substitution.lookup(p);
            if (arg == nullptr) {
                ::error("action %1%: parameter %2% must be bound", invocation, p);
                continue;
            }

            if (p->direction == IR::Direction::In ||
                p->direction == IR::Direction::InOut ||
                p->direction == IR::Direction::None) {
                auto left = new IR::PathExpression(p->name);
                auto assign = new IR::AssignmentStatement(arg->srcInfo, left, arg->expression);
                initializers->push_back(assign);
            }
            if (p->direction == IR::Direction::Out ||
                p->direction == IR::Direction::InOut) {
                auto right = new IR::PathExpression(p->name);
                auto assign = new IR::AssignmentStatement(arg->srcInfo, arg->expression, right);
                postamble->push_back(assign);
            }
        }
    }
    if (result->empty())
        return action;

    initializers->append(action->body->components);
    initializers->append(*postamble);

    action->parameters = new IR::ParameterList(action->parameters->srcInfo, *leftParams);
    action->body = new IR::BlockStatement(action->body->srcInfo, *initializers);
    LOG1("To replace " << dbp(action));
    result->push_back(action);
    return result;
}
Пример #3
0
bool
ErrorCodesVisitor::preorder(const IR::Type_Error* errors) {
    auto &map = *errorCodesMap;
    for (auto m : *errors->getDeclarations()) {
        BUG_CHECK(map.find(m) == map.end(), "Duplicate error");
        map[m] = map.size();
    }
    return false;
}
Пример #4
0
const IR::Node* DoConvertEnums::preorder(IR::Type_Enum* type) {
    bool convert = policy->convert(type);
    if (!convert)
        return type;
    unsigned long long count = type->members.size();
    unsigned long long width = policy->enumSize(count);
    LOG2("Converting enum " << type->name << " to " << "bit<" << width << ">");
    BUG_CHECK(count <= (1ULL << width),
              "%1%: not enough bits to represent %2%", width, type);
    auto r = new EnumRepresentation(type->srcInfo, width);
    auto canontype = typeMap->getTypeType(getOriginal(), true);
    BUG_CHECK(canontype->is<IR::Type_Enum>(),
              "canon type of enum %s is non enum %s?", type, canontype);
    repr.emplace(canontype->to<IR::Type_Enum>(), r);
    for (auto d : type->members)
        r->add(d->name.name);
    return nullptr;  // delete the declaration
}
Пример #5
0
/// Given an expression and a destination type, convert ListExpressions
/// that occur within expression to StructInitializerExpression if the
/// destination type matches.
const IR::Expression*
convert(const IR::Expression* expression, const IR::Type* type) {
    bool modified = false;
    if (auto st = type->to<IR::Type_StructLike>()) {
        auto si = new IR::IndexedVector<IR::NamedExpression>();
        if (auto le = expression->to<IR::ListExpression>()) {
            size_t index = 0;
            for (auto f : st->fields) {
                auto expr = le->components.at(index);
                auto conv = convert(expr, f->type);
                auto ne = new IR::NamedExpression(conv->srcInfo, f->name, conv);
                si->push_back(ne);
                index++;
            }
            auto result = new IR::StructInitializerExpression(
                expression->srcInfo, st->name, *si, st->is<IR::Type_Header>());
            return result;
        } else if (auto sli = expression->to<IR::StructInitializerExpression>()) {
            for (auto f : st->fields) {
                auto ne = sli->components.getDeclaration<IR::NamedExpression>(f->name.name);
                BUG_CHECK(ne != nullptr, "%1%: no initializer for %2%", expression, f);
                auto convNe = convert(ne->expression, f->type);
                if (convNe != ne->expression)
                    modified = true;
                ne = new IR::NamedExpression(ne->srcInfo, f->name, convNe);
                si->push_back(ne);
            }
            if (modified) {
                auto result = new IR::StructInitializerExpression(
                    expression->srcInfo, st->name, *si, st->is<IR::Type_Header>());
                return result;
            }
        }
    } else if (auto tup = type->to<IR::Type_Tuple>()) {
        auto le = expression->to<IR::ListExpression>();
        if (le == nullptr)
            return expression;

        auto vec = new IR::Vector<IR::Expression>();
        for (size_t i = 0; i < le->size(); i++) {
            auto expr = le->components.at(i);
            auto type = tup->components.at(i);
            auto conv = convert(expr, type);
            vec->push_back(conv);
            modified |= (conv != expr);
        }
        if (modified) {
            auto result = new IR::ListExpression(expression->srcInfo, *vec);
            return result;
        }
    }
    return expression;
}
Пример #6
0
void ControlBodyTranslator::compileEmit(const IR::Vector<IR::Argument>* args) {
    BUG_CHECK(args->size() == 1, "%1%: expected 1 argument for emit", args);

    auto expr = args->at(0)->expression;
    auto type = typeMap->getType(expr);
    auto ht = type->to<IR::Type_Header>();
    if (ht == nullptr) {
        ::error("Cannot emit a non-header type %1%", expr);
        return;
    }

    auto program = control->program;
    builder->emitIndent();
    builder->append("if (");
    visit(expr);
    builder->append(".ebpf_valid) ");
    builder->blockStart();

    unsigned width = ht->width_bits();
    builder->emitIndent();
    builder->appendFormat("if (%s < %s + BYTES(%s + %d)) ",
                          program->packetEndVar.c_str(),
                          program->packetStartVar.c_str(),
                          program->offsetVar.c_str(), width);
    builder->blockStart();

    builder->emitIndent();
    builder->appendFormat("%s = %s;", program->errorVar.c_str(),
                          p4lib.packetTooShort.str());
    builder->newline();

    builder->emitIndent();
    builder->appendFormat("return %s;", builder->target->abortReturnCode().c_str());
    builder->newline();
    builder->blockEnd(true);

    unsigned alignment = 0;
    for (auto f : ht->fields) {
        auto ftype = typeMap->getType(f);
        auto etype = EBPFTypeFactory::instance->create(ftype);
        auto et = dynamic_cast<IHasWidth*>(etype);
        if (et == nullptr) {
            ::error("Only headers with fixed widths supported %1%", f);
            return;
        }
        compileEmitField(expr, f->name, alignment, etype);
        alignment += et->widthInBits();
        alignment %= 8;
    }

    builder->blockEnd(true);
    return;
}
Пример #7
0
bool ControlBodyTranslator::preorder(const IR::SwitchStatement* statement) {
    cstring newName = control->program->refMap->newName("action_run");
    saveAction.push_back(newName);
    // This must be a table.apply().action_run
    auto mem = statement->expression->to<IR::Member>();
    BUG_CHECK(mem != nullptr,
              "%1%: Unexpected expression in switch statement", statement->expression);
    visit(mem->expr);
    saveAction.pop_back();
    saveAction.push_back(nullptr);
    builder->emitIndent();
    builder->append("switch (");
    builder->append(newName);
    builder->append(") ");
    builder->blockStart();
    for (auto c : statement->cases) {
        builder->emitIndent();
        if (c->label->is<IR::DefaultExpression>()) {
            builder->append("default");
        } else {
            builder->append("case ");
            auto pe = c->label->to<IR::PathExpression>();
            auto decl = control->program->refMap->getDeclaration(pe->path, true);
            BUG_CHECK(decl->is<IR::P4Action>(), "%1%: expected an action", pe);
            auto act = decl->to<IR::P4Action>();
            cstring name = EBPFObject::externalName(act);
            builder->append(name);
        }
        builder->append(":");
        builder->newline();
        builder->emitIndent();
        visit(c->statement);
        builder->newline();
        builder->emitIndent();
        builder->appendLine("break;");
    }
    builder->blockEnd(false);
    saveAction.pop_back();
    return false;
}
Пример #8
0
const IR::IDeclaration* ReferenceMap::getDeclaration(const IR::Path* path, bool notNull) const {
    CHECK_NULL(path);
    auto result = get(pathToDeclaration, path);

    if (result)
        LOG1("Looking up " << path << " found " << result->getNode());
    else
        LOG1("Looking up " << path << " found nothing");

    if (notNull)
        BUG_CHECK(result != nullptr, "Cannot find declaration for %1%", path);
    return result;
}
Пример #9
0
void DoCheckConstants::postorder(const IR::MethodCallExpression* expression) {
    auto mi = MethodInstance::resolve(expression, refMap, typeMap);
    if (auto bi = mi->to<BuiltInMethod>()) {
        if (bi->name == IR::Type_Stack::push_front ||
            bi->name == IR::Type_Stack::pop_front) {
            BUG_CHECK(expression->arguments->size() == 1,
                      "Expected 1 argument for %1%", expression);
            auto arg0 = expression->arguments->at(0)->expression;
            if (!arg0->is<IR::Constant>())
                ::error("%1%: argument must be a constant", arg0);
        }
    }
}
Пример #10
0
const IR::Node* TypeVariableSubstitutionVisitor::preorder(IR::TypeParameters *tps) {
    // remove all variables that were substituted
    for (auto it = tps->parameters.begin(); it != tps->parameters.end();) {
        const IR::Type* type = bindings->lookup(*it);
        if (type != nullptr && !replace) {
            LOG3("Removing from generic parameters " << *it);
            it = tps->parameters.erase(it);
        } else {
            if (type != nullptr)
                BUG_CHECK(type->is<IR::Type_Var>(),
                          "cannot replace a type parameter %1% with %2%:", *it, type);
            ++it;
        }
    }
    return tps;
}
Пример #11
0
void EBPFControl::emitDeclaration(CodeBuilder* builder, const IR::Declaration* decl) {
    if (decl->is<IR::Declaration_Variable>()) {
        auto vd = decl->to<IR::Declaration_Variable>();
        auto etype = EBPFTypeFactory::instance->create(vd->type);
        builder->emitIndent();
        etype->declare(builder, vd->name, false);
        builder->endOfStatement(true);
        BUG_CHECK(vd->initializer == nullptr,
                  "%1%: declarations with initializers not supported", decl);
        return;
    } else if (decl->is<IR::P4Table>() ||
               decl->is<IR::P4Action>() ||
               decl->is<IR::Declaration_Instance>()) {
        return;
    }
    BUG("%1%: not yet handled", decl);
}
Пример #12
0
const IR::StructInitializerExpression* StructTypeReplacement::explode(
    const IR::Expression *root, cstring prefix) {
    auto vec = new IR::IndexedVector<IR::NamedExpression>();
    auto fieldType = ::get(structFieldMap, prefix);
    BUG_CHECK(fieldType, "No field for %1%", prefix);
    for (auto f : fieldType->fields) {
        cstring fieldName = prefix + "." + f->name.name;
        auto newFieldname = ::get(fieldNameRemap, fieldName);
        const IR::Expression* expr;
        if (!newFieldname.isNullOrEmpty()) {
            expr = new IR::Member(root, newFieldname);
        } else {
            expr = explode(root, fieldName);
        }
        vec->push_back(new IR::NamedExpression(f->name, expr));
    }
    return new IR::StructInitializerExpression(fieldType->name, *vec, false);
}
Пример #13
0
const IR::Node* DoResetHeaders::postorder(IR::Declaration_Variable* decl) {
    if (findContext<IR::ParserState>() == nullptr)
        return decl;
    if (decl->initializer != nullptr)
        return decl;
    auto resets = new IR::Vector<IR::StatOrDecl>();
    resets->push_back(decl);
    BUG_CHECK(getContext()->node->is<IR::Vector<IR::StatOrDecl>>() ||
              getContext()->node->is<IR::ParserState>() ||
              getContext()->node->is<IR::BlockStatement>(),
              "%1%: parent is not Vector<StatOrDecl>, but %2%",
              decl, getContext()->node);
    auto type = typeMap->getType(getOriginal(), true);
    auto path = new IR::PathExpression(decl->getName());
    generateResets(typeMap, type, path, resets);
    if (resets->size() == 1)
        return decl;
    return resets;
}
Пример #14
0
const IR::Node* DoBindTypeVariables::postorder(IR::MethodCallExpression* expression) {
    if (!expression->typeArguments->empty())
        return expression;
    auto type = typeMap->getType(expression->method, true);
    BUG_CHECK(type->is<IR::IMayBeGenericType>(), "%1%: unexpected type %2% for method",
              expression->method, type);
    auto mt = type->to<IR::IMayBeGenericType>();
    if (mt->getTypeParameters()->empty())
        return expression;
    auto typeArgs = new IR::Vector<IR::Type>();
    for (auto p : mt->getTypeParameters()->parameters) {
        auto type = getVarValue(p, expression);
        if (type == nullptr)
            return expression;
        typeArgs->push_back(type);
    }
    expression->typeArguments = typeArgs;
    return expression;
}
Пример #15
0
const IR::Node* DoBindTypeVariables::postorder(IR::Declaration_Instance* decl) {
    if (decl->type->is<IR::Type_Specialized>())
        return decl;
    auto type = typeMap->getType(getOriginal(), true);
    BUG_CHECK(type->is<IR::IMayBeGenericType>(), "%1%: unexpected type %2% for declaration",
              decl, type);
    auto mt = type->to<IR::IMayBeGenericType>();
    if (mt->getTypeParameters()->empty())
        return decl;
    auto typeArgs = new IR::Vector<IR::Type>();
    for (auto p : mt->getTypeParameters()->parameters) {
        auto type = getVarValue(p, decl);
        if (type == nullptr)
            return decl;
        typeArgs->push_back(type);
    }
    decl->type = new IR::Type_Specialized(
        decl->type->srcInfo, decl->type->to<IR::Type_Name>(), typeArgs);
    return decl;
}
Пример #16
0
const IR::Node* DoBindTypeVariables::postorder(IR::ConstructorCallExpression* expression) {
    if (expression->constructedType->is<IR::Type_Specialized>())
        return expression;
    auto type = typeMap->getType(getOriginal(), true);
    BUG_CHECK(type->is<IR::IMayBeGenericType>(), "%1%: unexpected type %2% for expression",
              expression, type);
    auto mt = type->to<IR::IMayBeGenericType>();
    if (mt->getTypeParameters()->empty())
        return expression;
    auto typeArgs = new IR::Vector<IR::Type>();
    for (auto p : mt->getTypeParameters()->parameters) {
        auto type = getVarValue(p, expression);
        if (type == nullptr)
            return expression;
        typeArgs->push_back(type);
    }
    expression->constructedType = new IR::Type_Specialized(
        expression->constructedType->srcInfo,
        expression->constructedType->to<IR::Type_Name>(), typeArgs);
    return expression;
}
Пример #17
0
clfftStatus FFTAction::enqueue(clfftPlanHandle plHandle,
                               clfftDirection dir,
                               cl_uint numQueuesAndEvents,
                               cl_command_queue* commQueues,
                               cl_uint numWaitEvents,
                               const cl_event* waitEvents,
                               cl_event* outEvents,
                               cl_mem* clInputBuffers,
                               cl_mem* clOutputBuffers)
{
    FFTRepo & fftRepo = FFTRepo::getInstance();

    std::vector< cl_mem > inputBuff;
    std::vector< cl_mem > outputBuff;


    clfftStatus status = selectBufferArguments(this->plan,
                                               clInputBuffers, clOutputBuffers,
                                               inputBuff, outputBuff);

    if (status != CLFFT_SUCCESS)
    {
        return status;
    }
    
    //	TODO:  In the case of length == 1, FFT is a trivial NOP, but we still need to apply the forward and backwards tranforms
    //	TODO:  Are map lookups expensive to call here?  We can cache a pointer to the cl_program/cl_kernel in the plan

    //	Translate the user plan into the structure that we use to map plans to clPrograms

    cl_program	prog;
    cl_kernel	kern;
    OPENCL_V( fftRepo.getclProgram( this->getGenerator(), this->getSignatureData(), prog, this->plan->bakeDevice, this->plan->context ), _T( "fftRepo.getclProgram failed" ) );
    OPENCL_V( fftRepo.getclKernel( prog, dir, kern ), _T( "fftRepo.getclKernels failed" ) );

    cl_uint uarg = 0;
    if (!this->plan->transflag && !(this->plan->gen == Copy))
    {
        //	::clSetKernelArg() is not thread safe, according to the openCL spec for the same cl_kernel object
        //	TODO:  Need to verify that two different plans (which would get through our lock above) with exactly the same
        //	parameters would NOT share the same cl_kernel objects

        /* constant buffer */
        OPENCL_V( clSetKernelArg( kern, uarg++, sizeof( cl_mem ), (void*)&this->plan->const_buffer ), _T( "clSetKernelArg failed" ) );
    }

    //	Input buffer(s)
    //	Input may be 1 buffer  (CLFFT_COMPLEX_INTERLEAVED)
    //	          or 2 buffers (CLFFT_COMPLEX_PLANAR)

    for (size_t i = 0; i < inputBuff.size(); ++i)
    {
        OPENCL_V( clSetKernelArg( kern, uarg++, sizeof( cl_mem ), (void*)&inputBuff[i] ), _T( "clSetKernelArg failed" ) );
    }
    //	Output buffer(s)
    //	Output may be 0 buffers (CLFFT_INPLACE)
    //	           or 1 buffer  (CLFFT_COMPLEX_INTERLEAVED)
    //	           or 2 buffers (CLFFT_COMPLEX_PLANAR)
    for (size_t o = 0; o < outputBuff.size(); ++o)
    {
        OPENCL_V( clSetKernelArg( kern, uarg++, sizeof( cl_mem ), (void*)&outputBuff[o] ), _T( "clSetKernelArg failed" ) );
    }

	//If callback function is set for the plan, pass the appropriate aruments
	if (this->plan->hasPreCallback || this->plan->hasPostCallback)
	{
	if (this->plan->hasPreCallback)
	{
		OPENCL_V( clSetKernelArg( kern, uarg++, sizeof( cl_mem ), (void*)&this->plan->precallUserData ), _T( "clSetKernelArg failed" ) );
		}

		//If post-callback function is set for the plan, pass the appropriate aruments
		if (this->plan->hasPostCallback)
		{
			OPENCL_V( clSetKernelArg( kern, uarg++, sizeof( cl_mem ), (void*)&this->plan->postcallUserData ), _T( "clSetKernelArg failed" ) );
		}

		//Pass LDS size arument if set
		if ((this->plan->hasPreCallback && this->plan->preCallback.localMemSize > 0) || 
			(this->plan->hasPostCallback && this->plan->postCallbackParam.localMemSize > 0))
		{
			int localmemSize = 0;
			if (this->plan->hasPreCallback && this->plan->preCallback.localMemSize > 0)
				localmemSize = this->plan->preCallback.localMemSize;
			if (this->plan->hasPostCallback && this->plan->postCallbackParam.localMemSize > 0)
				localmemSize += this->plan->postCallbackParam.localMemSize;

			OPENCL_V( clSetKernelArg( kern, uarg++, localmemSize, NULL ), _T( "clSetKernelArg failed" ) );
		}
	}

    std::vector< size_t > gWorkSize;
    std::vector< size_t > lWorkSize;
    clfftStatus result = this->getWorkSizes (gWorkSize, lWorkSize);

    // TODO:  if getWorkSizes returns CLFFT_INVALID_GLOBAL_WORK_SIZE, that means
    // that this multidimensional input data array is too large to be transformed
    // with a single call to clEnqueueNDRangeKernel.  For now, we will just return
    // the error code back up the call stack.
    // The *correct* course of action would be to split the work into mutliple
    // calls to clEnqueueNDRangeKernel.
    if (CLFFT_INVALID_GLOBAL_WORK_SIZE == result)
    {
        OPENCL_V( result, _T("Work size too large for clEnqueNDRangeKernel()"));
    }
    else
    {
        OPENCL_V( result, _T("FFTAction::getWorkSizes failed"));
    }
    BUG_CHECK (gWorkSize.size() == lWorkSize.size());


    cl_int call_status = clEnqueueNDRangeKernel( *commQueues, kern, static_cast< cl_uint >( gWorkSize.size( ) ),
                                            NULL, &gWorkSize[ 0 ],  &lWorkSize[ 0 ], numWaitEvents, waitEvents, outEvents );
    OPENCL_V( call_status, _T( "clEnqueueNDRangeKernel failed" ) );

    if( fftRepo.pStatTimer )
    {
        fftRepo.pStatTimer->AddSample( plHandle, this->plan, kern, numQueuesAndEvents, outEvents, gWorkSize, lWorkSize );
    }

    return CLFFT_SUCCESS;
}
Пример #18
0
void ControlBodyTranslator::processApply(const P4::ApplyMethod* method) {
    builder->emitIndent();
    auto table = control->getTable(method->object->getName().name);
    BUG_CHECK(table != nullptr, "No table for %1%", method->expr);

    P4::ParameterSubstitution binding;
    cstring actionVariableName;
    if (!saveAction.empty()) {
        actionVariableName = saveAction.at(saveAction.size() - 1);
        if (!actionVariableName.isNullOrEmpty()) {
            builder->appendFormat("enum %s %s;\n",
                                  table->actionEnumName.c_str(), actionVariableName.c_str());
            builder->emitIndent();
        }
    }
    builder->blockStart();

    BUG_CHECK(method->expr->arguments->size() == 0, "%1%: table apply with arguments", method);
    cstring keyname = "key";
    if (table->keyGenerator != nullptr) {
        builder->emitIndent();
        builder->appendLine("/* construct key */");
        builder->emitIndent();
        builder->appendFormat("struct %s %s = {}", table->keyTypeName.c_str(), keyname.c_str());
        builder->endOfStatement(true);
        table->emitKey(builder, keyname);
    }
    builder->emitIndent();
    builder->appendLine("/* value */");
    builder->emitIndent();
    cstring valueName = "value";
    builder->appendFormat("struct %s *%s = NULL", table->valueTypeName.c_str(), valueName.c_str());
    builder->endOfStatement(true);

    if (table->keyGenerator != nullptr) {
        builder->emitIndent();
        builder->appendLine("/* perform lookup */");
        builder->emitIndent();
        builder->target->emitTableLookup(builder, table->dataMapName, keyname, valueName);
        builder->endOfStatement(true);
    }

    builder->emitIndent();
    builder->appendFormat("if (%s == NULL) ", valueName.c_str());
    builder->blockStart();

    builder->emitIndent();
    builder->appendLine("/* miss; find default action */");
    builder->emitIndent();
    builder->appendFormat("%s = 0", control->hitVariable.c_str());
    builder->endOfStatement(true);

    builder->emitIndent();
    builder->target->emitTableLookup(builder, table->defaultActionMapName,
                                     control->program->zeroKey, valueName);
    builder->endOfStatement(true);
    builder->blockEnd(false);
    builder->append(" else ");
    builder->blockStart();
    builder->emitIndent();
    builder->appendFormat("%s = 1", control->hitVariable.c_str());
    builder->endOfStatement(true);
    builder->blockEnd(true);

    builder->emitIndent();
    builder->appendFormat("if (%s != NULL) ", valueName.c_str());
    builder->blockStart();
    builder->emitIndent();
    builder->appendLine("/* run action */");
    table->emitAction(builder, valueName);
    if (!actionVariableName.isNullOrEmpty()) {
        builder->emitIndent();
        builder->appendFormat("%s = %s->action",
                              actionVariableName.c_str(), valueName.c_str());
        builder->endOfStatement(true);
    }
    toDereference.clear();

    builder->blockEnd(true);
    builder->emitIndent();
    builder->appendFormat("else return %s", builder->target->abortReturnCode().c_str());
    builder->endOfStatement(true);

    builder->blockEnd(true);
}
Пример #19
0
void ControlBodyTranslator::compileEmitField(const IR::Expression* expr, cstring field,
                                             unsigned alignment, EBPFType* type) {
    unsigned widthToEmit = dynamic_cast<IHasWidth*>(type)->widthInBits();
    cstring swap = "";
    if (widthToEmit == 16)
        swap = "htons";
    else if (widthToEmit == 32)
        swap = "htonl";
    if (!swap.isNullOrEmpty()) {
        builder->emitIndent();
        visit(expr);
        builder->appendFormat(".%s = %s(", field.c_str(), swap);
        visit(expr);
        builder->appendFormat(".%s)", field.c_str());
        builder->endOfStatement(true);
    }

    auto program = control->program;
    unsigned bitsInFirstByte = widthToEmit % 8;
    if (bitsInFirstByte == 0) bitsInFirstByte = 8;
    unsigned bitsInCurrentByte = bitsInFirstByte;
    unsigned left = widthToEmit;

    for (unsigned i=0; i < (widthToEmit + 7) / 8; i++) {
        builder->emitIndent();
        builder->appendFormat("%s = ((char*)(&", program->byteVar.c_str());
        visit(expr);
        builder->appendFormat(".%s))[%d]", field.c_str(), i);
        builder->endOfStatement(true);

        unsigned freeBits = alignment == 0 ? (8 - alignment) : 8;
        unsigned bitsToWrite = bitsInCurrentByte > freeBits ? freeBits : bitsInCurrentByte;

        BUG_CHECK((bitsToWrite > 0) && (bitsToWrite <= 8), "invalid bitsToWrite %d", bitsToWrite);
        builder->emitIndent();
        if (alignment == 0)
            builder->appendFormat("write_byte(%s, BYTES(%s) + %d, (%s) << %d)",
                                  program->packetStartVar.c_str(), program->offsetVar.c_str(), i,
                                  program->byteVar.c_str(), 8 - bitsToWrite);
        else
            builder->appendFormat("write_partial(%s + BYTES(%s) + %d, %d, (%s) << %d)",
                                  program->packetStartVar.c_str(), program->offsetVar.c_str(), i,
                                  alignment,
                                  program->byteVar.c_str(), 8 - bitsToWrite);
        builder->endOfStatement(true);
        left -= bitsToWrite;
        bitsInCurrentByte -= bitsToWrite;

        if (bitsInCurrentByte > 0) {
            builder->emitIndent();
            builder->appendFormat(
                "write_byte(%s, BYTES(%s) + %d + 1, (%s << %d))",
                program->packetStartVar.c_str(),
                program->offsetVar.c_str(), i, program->byteVar.c_str(), 8 - alignment % 8);
            builder->endOfStatement(true);
            left -= bitsInCurrentByte;
        }

        alignment = (alignment + bitsToWrite) % 8;
        bitsInCurrentByte = left >= 8 ? 8 : left;
    }

    builder->emitIndent();
    builder->appendFormat("%s += %d", program->offsetVar.c_str(), widthToEmit);
    builder->endOfStatement(true);
}
Пример #20
0
bool ControlBodyTranslator::preorder(const IR::MethodCallExpression* expression) {
    builder->append("/* ");
    visit(expression->method);
    builder->append("(");
    bool first = true;
    for (auto a  : *expression->arguments) {
        if (!first)
            builder->append(", ");
        first = false;
        visit(a);
    }
    builder->append(")");
    builder->append("*/");
    builder->newline();

    auto mi = P4::MethodInstance::resolve(expression,
                                          control->program->refMap,
                                          control->program->typeMap);
    auto apply = mi->to<P4::ApplyMethod>();
    if (apply != nullptr) {
        processApply(apply);
        return false;
    }
    auto ef = mi->to<P4::ExternFunction>();
    if (ef != nullptr) {
        processFunction(ef);
        return false;
    }
    auto ext = mi->to<P4::ExternMethod>();
    if (ext != nullptr) {
        processMethod(ext);
        return false;
    }
    auto bim = mi->to<P4::BuiltInMethod>();
    if (bim != nullptr) {
        builder->emitIndent();
        if (bim->name == IR::Type_Header::isValid) {
            visit(bim->appliedTo);
            builder->append(".ebpf_valid");
            return false;
        } else if (bim->name == IR::Type_Header::setValid) {
            visit(bim->appliedTo);
            builder->append(".ebpf_valid = true");
            return false;
        } else if (bim->name == IR::Type_Header::setInvalid) {
            visit(bim->appliedTo);
            builder->append(".ebpf_valid = false");
            return false;
        }
    }
    auto ac = mi->to<P4::ActionCall>();
    if (ac != nullptr) {
        // Action arguments have been eliminated by the mid-end.
        BUG_CHECK(expression->arguments->size() == 0,
                  "%1%: unexpected arguments for action call", expression);
        visit(ac->action->body);
        return false;
    }

    ::error("Unsupported method invocation %1%", expression);
    return false;
}
Пример #21
0
const IR::Node* DoMoveActionsToTables::postorder(IR::MethodCallStatement* statement) {
    auto mi = MethodInstance::resolve(statement, refMap, typeMap);
    if (!mi->is<ActionCall>())
        return statement;
    auto ac = mi->to<ActionCall>();

    auto action = ac->action;
    auto directionArgs = new IR::Vector<IR::Argument>();
    auto mc = statement->methodCall;

    // TODO: should use argument names
    auto it = action->parameters->parameters.begin();
    auto arg = mc->arguments->begin();
    for (; it != action->parameters->parameters.end(); ++it) {
        auto p = *it;
        if (p->direction == IR::Direction::None)
            break;
        directionArgs->push_back(*arg);
        ++arg;
    }

    // Action invocation
    BUG_CHECK(ac->expr->method->is<IR::PathExpression>(),
              "%1%: Expected a PathExpression", ac->expr->method);
    auto actionPath = new IR::PathExpression(IR::ID(mc->srcInfo, ac->action->name));
    auto call = new IR::MethodCallExpression(mc->srcInfo, actionPath,
                                             new IR::Vector<IR::Type>(), directionArgs);
    auto actinst = new IR::ActionListElement(statement->srcInfo, call);
    // Action list property
    auto actlist = new IR::ActionList({actinst});
    auto prop = new IR::Property(
        IR::ID(IR::TableProperties::actionsPropertyName, nullptr),
        actlist, false);
    // default action property
    auto otherArgs = new IR::Vector<IR::Argument>();
    for (; it != action->parameters->parameters.end(); ++it) {
        otherArgs->push_back(*arg);
        ++arg;
    }
    BUG_CHECK(arg == mc->arguments->end(), "%1%: mismatched arguments", mc);
    auto amce = new IR::MethodCallExpression(mc->srcInfo, mc->method, mc->typeArguments, otherArgs);
    auto defactval = new IR::ExpressionValue(amce);
    auto defprop = new IR::Property(
        IR::ID(IR::TableProperties::defaultActionPropertyName, nullptr),
        defactval, true);

    // List of table properties
    auto props = new IR::TableProperties({ prop, defprop });
    // Synthesize a new table
    cstring tblName = IR::ID(refMap->newName(cstring("tbl_") + ac->action->name.name), nullptr);

    auto annos = new IR::Annotations();
    annos->add(new IR::Annotation(IR::Annotation::hiddenAnnotation, {}));
    auto tbl = new IR::P4Table(tblName, annos, props);
    tables.push_back(tbl);

    // Table invocation statement
    auto tblpath = new IR::PathExpression(tblName);
    auto method = new IR::Member(tblpath, IR::IApply::applyMethodName);
    auto mce = new IR::MethodCallExpression(
        statement->srcInfo, method, new IR::Vector<IR::Type>(),
        new IR::Vector<IR::Argument>());
    auto stat = new IR::MethodCallStatement(mce->srcInfo, mce);
    return stat;
}