8 #include "WTADelivery.hpp" 9 #include "columns/HyPerCol.hpp" 10 #include "utils/MapLookupByType.hpp" 15 WTADelivery::WTADelivery(
char const *name, HyPerCol *hc) { initialize(name, hc); }
17 int WTADelivery::initialize(
char const *name, HyPerCol *hc) {
18 return BaseDelivery::initialize(name, hc);
21 void WTADelivery::setObjectType() { mObjectType =
"WTADelivery"; }
26 if (ioFlag == PARAMS_IO_READ) {
27 parent->parameters()->handleUnnecessaryParameter(name,
"receiveGpu",
false );
32 WTADelivery::communicateInitInfo(std::shared_ptr<CommunicateInitInfoMessage const> message) {
33 auto status = BaseDelivery::communicateInitInfo(message);
34 if (status != Response::SUCCESS) {
38 auto *singleArbor = mapLookupByType<SingleArbor>(message->mHierarchy, getDescription());
39 FatalIf(!singleArbor,
"%s requires a SingleArbor component.\n", getDescription_c());
40 if (!singleArbor->getInitInfoCommunicatedFlag()) {
41 return Response::POSTPONE;
45 return Response::SUCCESS;
49 int status = PV_SUCCESS;
50 pvAssert(mPreLayer and mPostLayer);
51 PVLayerLoc const *preLoc = mPreLayer->getLayerLoc();
52 PVLayerLoc const *postLoc = mPostLayer->getLayerLoc();
53 if (preLoc->nx != postLoc->nx) {
55 "%s requires pre and post nx be equal (%d versus %d).\n",
61 if (preLoc->ny != postLoc->ny) {
63 "%s requires pre and post ny be equal (%d versus %d).\n",
69 if (postLoc->nf != 1) {
71 "%s requires post nf be equal to 1 (observed value is %d).\n",
76 if (preLoc->nbatch != postLoc->nbatch) {
78 "%s requires pre and post nbatch be equal (%d versus %d).\n",
86 "WTADelivery \"%s\" Error: %s and %s do not have the same dimensions.\n Dims: " 87 "%dx%dx%d vs. %dx%dx%d\n",
90 mPostLayer->getName(),
99 void WTADelivery::deliver() {
100 if (mChannelCode == CHANNEL_NOUPDATE) {
105 PVLayerLoc const &preLoc = preActivityCube.loc;
107 int const nx = preLoc.nx;
108 int const ny = preLoc.ny;
109 int const nf = preLoc.nf;
110 PVHalo const &halo = preLoc.halo;
111 int nxPreExtended = nx + halo.lt + halo.rt;
112 int nyPreExtended = ny + halo.dn + halo.up;
113 int numPreExtended = nxPreExtended * nyPreExtended * nf;
114 pvAssert(numPreExtended * preLoc.nbatch == preActivityCube.numItems);
115 int numPostRestricted = nx * ny * nf;
117 float *postChannel = mPostLayer->getChannel(mChannelCode);
118 for (
int b = 0; b < parent->getNBatch(); b++) {
119 float const *preActivityBuffer = preActivityCube.data + b * numPreExtended;
120 float *postGSynBuffer = postChannel + b * numPostRestricted;
127 int const numLocations = nx * ny;
128 #ifdef PV_USE_OPENMP_THREADS 129 #pragma omp parallel for 131 for (
int k = 0; k < numLocations; k++) {
132 int const kPreExtended =
133 kIndexExtended(k, nx, ny, 1, halo.lt, halo.rt, halo.dn, halo.up) * nf;
134 float maxValue = -FLT_MAX;
135 for (
int f = 0; f < nf; f++) {
136 float value = preActivityBuffer[kPreExtended + f];
137 maxValue = value > maxValue ? value : maxValue;
139 postGSynBuffer[k] += maxValue;
143 mPostLayer->setUpdatedDeviceGSynFlag(!mReceiveGpu);
144 #endif // PV_USE_CUDA 147 void WTADelivery::deliverUnitInput(
float *recvBuffer) {
148 const int numNeuronsPost = mPostLayer->getNumNeuronsAllBatches();
149 #ifdef PV_USE_OPENMP_THREADS 150 #pragma omp parallel for 152 for (
int k = 0; k < numNeuronsPost; k++) {
153 recvBuffer[k] += 1.0f;
159 if (getChannelCode() == CHANNEL_NOUPDATE) {
PVLayerCube createCube(int delay=0)
bool isExchangeFinished(int delay=0)
virtual bool isAllInputReady() override
virtual void ioParam_receiveGpu(enum ParamsIOFlag ioFlag) override
WTADelivery does not use the GPU. It is an error to set receiveGpu to true.
void checkPreAndPostDimensions()