src/gallium/drivers/swr/rasterizer/core/tilemgr.cpp

   1 /****************************************************************************
   2 * Copyright (C) 2014-2015 Intel Corporation.   All Rights Reserved.
   3 *
   4 * Permission is hereby granted, free of charge, to any person obtaining a
   5 * copy of this software and associated documentation files (the "Software"),
   6 * to deal in the Software without restriction, including without limitation
   7 * the rights to use, copy, modify, merge, publish, distribute, sublicense,
   8 * and/or sell copies of the Software, and to permit persons to whom the
   9 * Software is furnished to do so, subject to the following conditions:
  10 *
  11 * The above copyright notice and this permission notice (including the next
  12 * paragraph) shall be included in all copies or substantial portions of the
  13 * Software.
  14 *
  15 * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
  16 * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
  17 * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
  18 * THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
  19 * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING
  20 * FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS
  21 * IN THE SOFTWARE.
  22 *
  23 * @file tilemgr.cpp
  24 *
  25 * @brief Implementation for Macro Tile Manager which provides the facilities
  26 *        for threads to work on an macro tile.
  27 *
  28 ******************************************************************************/
  29 #include <unordered_map>
  30
  31 #include "fifo.hpp"
  32 #include "core/tilemgr.h"
  33 #include "core/multisample.h"
  34 #include "rdtsc_core.h"
  35
  36 #define TILE_ID(x,y) ((x << 16 | y))
  37
  38 MacroTileMgr::MacroTileMgr(CachingArena& arena) : mArena(arena)
  39 {
  40 }
  41
  42 void MacroTileMgr::enqueue(uint32_t x, uint32_t y, BE_WORK *pWork)
  43 {
  44     // Should not enqueue more then what we have backing for in the hot tile manager.
  45     SWR_ASSERT(x < KNOB_NUM_HOT_TILES_X);
  46     SWR_ASSERT(y < KNOB_NUM_HOT_TILES_Y);
  47
  48     if ((x & ~(KNOB_NUM_HOT_TILES_X-1)) | (y & ~(KNOB_NUM_HOT_TILES_Y-1)))
  49     {
  50         return;
  51     }
  52
  53     uint32_t id = TILE_ID(x, y);
  54
  55     MacroTileQueue &tile = mTiles[id];
  56     tile.mWorkItemsFE++;
  57     tile.mId = id;
  58
  59     if (tile.mWorkItemsFE == 1)
  60     {
  61         tile.clear(mArena);
  62         mDirtyTiles.push_back(&tile);
  63     }
  64
  65     mWorkItemsProduced++;
  66     tile.enqueue_try_nosync(mArena, pWork);
  67 }
  68
  69 void MacroTileMgr::markTileComplete(uint32_t id)
  70 {
  71     SWR_ASSERT(mTiles.find(id) != mTiles.end());
  72     MacroTileQueue &tile = mTiles[id];
  73     uint32_t numTiles = tile.mWorkItemsFE;
  74     InterlockedExchangeAdd(&mWorkItemsConsumed, numTiles);
  75
  76     _ReadWriteBarrier();
  77     tile.mWorkItemsBE += numTiles;
  78     SWR_ASSERT(tile.mWorkItemsFE == tile.mWorkItemsBE);
  79
  80     // clear out tile, but defer fifo clear until the next DC first queues to it.
  81     // this prevents worker threads from constantly locking a completed macro tile
  82     tile.mWorkItemsFE = 0;
  83     tile.mWorkItemsBE = 0;
  84 }
  85
  86 HOTTILE* HotTileMgr::GetHotTile(SWR_CONTEXT* pContext, DRAW_CONTEXT* pDC, uint32_t macroID, SWR_RENDERTARGET_ATTACHMENT attachment, bool create, uint32_t numSamples,
  87     uint32_t renderTargetArrayIndex)
  88 {
  89     uint32_t x, y;
  90     MacroTileMgr::getTileIndices(macroID, x, y);
  91
  92     SWR_ASSERT(x < KNOB_NUM_HOT_TILES_X);
  93     SWR_ASSERT(y < KNOB_NUM_HOT_TILES_Y);
  94
  95     HotTileSet &tile = mHotTiles[x][y];
  96     HOTTILE& hotTile = tile.Attachment[attachment];
  97     if (hotTile.pBuffer == NULL)
  98     {
  99         if (create)
 100         {
 101             uint32_t size = numSamples * mHotTileSize[attachment];
 102             uint32_t numaNode = ((x ^ y) & pContext->threadPool.numaMask);
 103             hotTile.pBuffer = (uint8_t*)AllocHotTileMem(size, KNOB_SIMD_WIDTH * 4, numaNode);
 104             hotTile.state = HOTTILE_INVALID;
 105             hotTile.numSamples = numSamples;
 106             hotTile.renderTargetArrayIndex = renderTargetArrayIndex;
 107         }
 108         else
 109         {
 110             return NULL;
 111         }
 112     }
 113     else
 114     {
 115         // free the old tile and create a new one with enough space to hold all samples
 116         if (numSamples > hotTile.numSamples)
 117         {
 118             // tile should be either uninitialized or resolved if we're deleting and switching to a
 119             // new sample count
 120             SWR_ASSERT((hotTile.state == HOTTILE_INVALID) ||
 121                 (hotTile.state == HOTTILE_RESOLVED) ||
 122                 (hotTile.state == HOTTILE_CLEAR));
 123             FreeHotTileMem(hotTile.pBuffer);
 124
 125             uint32_t size = numSamples * mHotTileSize[attachment];
 126             uint32_t numaNode = ((x ^ y) & pContext->threadPool.numaMask);
 127             hotTile.pBuffer = (uint8_t*)AllocHotTileMem(size, KNOB_SIMD_WIDTH * 4, numaNode);
 128             hotTile.state = HOTTILE_INVALID;
 129             hotTile.numSamples = numSamples;
 130         }
 131
 132         // if requested render target array index isn't currently loaded, need to store out the current hottile
 133         // and load the requested array slice
 134         if (renderTargetArrayIndex != hotTile.renderTargetArrayIndex)
 135         {
 136             SWR_FORMAT format;
 137             switch (attachment)
 138             {
 139             case SWR_ATTACHMENT_COLOR0:
 140             case SWR_ATTACHMENT_COLOR1:
 141             case SWR_ATTACHMENT_COLOR2:
 142             case SWR_ATTACHMENT_COLOR3:
 143             case SWR_ATTACHMENT_COLOR4:
 144             case SWR_ATTACHMENT_COLOR5:
 145             case SWR_ATTACHMENT_COLOR6:
 146             case SWR_ATTACHMENT_COLOR7: format = KNOB_COLOR_HOT_TILE_FORMAT; break;
 147             case SWR_ATTACHMENT_DEPTH: format = KNOB_DEPTH_HOT_TILE_FORMAT; break;
 148             case SWR_ATTACHMENT_STENCIL: format = KNOB_STENCIL_HOT_TILE_FORMAT; break;
 149             default: SWR_ASSERT(false, "Unknown attachment: %d", attachment); format = KNOB_COLOR_HOT_TILE_FORMAT; break;
 150             }
 151
 152             if (hotTile.state == HOTTILE_DIRTY)
 153             {
 154                 pContext->pfnStoreTile(GetPrivateState(pDC), format, attachment,
 155                     x * KNOB_MACROTILE_X_DIM, y * KNOB_MACROTILE_Y_DIM, hotTile.renderTargetArrayIndex, hotTile.pBuffer);
 156             }
 157
 158             pContext->pfnLoadTile(GetPrivateState(pDC), format, attachment,
 159                 x * KNOB_MACROTILE_X_DIM, y * KNOB_MACROTILE_Y_DIM, renderTargetArrayIndex, hotTile.pBuffer);
 160
 161             hotTile.renderTargetArrayIndex = renderTargetArrayIndex;
 162             hotTile.state = HOTTILE_DIRTY;
 163         }
 164     }
 165     return &tile.Attachment[attachment];
 166 }
 167
 168 HOTTILE* HotTileMgr::GetHotTileNoLoad(
 169     SWR_CONTEXT* pContext, DRAW_CONTEXT* pDC, uint32_t macroID,
 170     SWR_RENDERTARGET_ATTACHMENT attachment, bool create, uint32_t numSamples)
 171 {
 172     uint32_t x, y;
 173     MacroTileMgr::getTileIndices(macroID, x, y);
 174
 175     SWR_ASSERT(x < KNOB_NUM_HOT_TILES_X);
 176     SWR_ASSERT(y < KNOB_NUM_HOT_TILES_Y);
 177
 178     HotTileSet &tile = mHotTiles[x][y];
 179     HOTTILE& hotTile = tile.Attachment[attachment];
 180     if (hotTile.pBuffer == NULL)
 181     {
 182         if (create)
 183         {
 184             uint32_t size = numSamples * mHotTileSize[attachment];
 185             hotTile.pBuffer = (uint8_t*)AlignedMalloc(size, KNOB_SIMD_WIDTH * 4);
 186             hotTile.state = HOTTILE_INVALID;
 187             hotTile.numSamples = numSamples;
 188             hotTile.renderTargetArrayIndex = 0;
 189         }
 190         else
 191         {
 192             return NULL;
 193         }
 194     }
 195
 196     return &hotTile;
 197 }
 198
 199 void HotTileMgr::ClearColorHotTile(const HOTTILE* pHotTile)  // clear a macro tile from float4 clear data.
 200 {
 201     // Load clear color into SIMD register...
 202     float *pClearData = (float*)(pHotTile->clearData);
 203     simdscalar valR = _simd_broadcast_ss(&pClearData[0]);
 204     simdscalar valG = _simd_broadcast_ss(&pClearData[1]);
 205     simdscalar valB = _simd_broadcast_ss(&pClearData[2]);
 206     simdscalar valA = _simd_broadcast_ss(&pClearData[3]);
 207
 208     float *pfBuf = (float*)pHotTile->pBuffer;
 209     uint32_t numSamples = pHotTile->numSamples;
 210
 211     for (uint32_t row = 0; row < KNOB_MACROTILE_Y_DIM; row += KNOB_TILE_Y_DIM)
 212     {
 213         for (uint32_t col = 0; col < KNOB_MACROTILE_X_DIM; col += KNOB_TILE_X_DIM)
 214         {
 215             for (uint32_t si = 0; si < (KNOB_TILE_X_DIM * KNOB_TILE_Y_DIM * numSamples); si += SIMD_TILE_X_DIM * SIMD_TILE_Y_DIM) //SIMD_TILE_X_DIM * SIMD_TILE_Y_DIM); si++)
 216             {
 217                 _simd_store_ps(pfBuf, valR);
 218                 pfBuf += KNOB_SIMD_WIDTH;
 219                 _simd_store_ps(pfBuf, valG);
 220                 pfBuf += KNOB_SIMD_WIDTH;
 221                 _simd_store_ps(pfBuf, valB);
 222                 pfBuf += KNOB_SIMD_WIDTH;
 223                 _simd_store_ps(pfBuf, valA);
 224                 pfBuf += KNOB_SIMD_WIDTH;
 225             }
 226         }
 227     }
 228 }
 229
 230 void HotTileMgr::ClearDepthHotTile(const HOTTILE* pHotTile)  // clear a macro tile from float4 clear data.
 231 {
 232     // Load clear color into SIMD register...
 233     float *pClearData = (float*)(pHotTile->clearData);
 234     simdscalar valZ = _simd_broadcast_ss(&pClearData[0]);
 235
 236     float *pfBuf = (float*)pHotTile->pBuffer;
 237     uint32_t numSamples = pHotTile->numSamples;
 238
 239     for (uint32_t row = 0; row < KNOB_MACROTILE_Y_DIM; row += KNOB_TILE_Y_DIM)
 240     {
 241         for (uint32_t col = 0; col < KNOB_MACROTILE_X_DIM; col += KNOB_TILE_X_DIM)
 242         {
 243             for (uint32_t si = 0; si < (KNOB_TILE_X_DIM * KNOB_TILE_Y_DIM * numSamples); si += SIMD_TILE_X_DIM * SIMD_TILE_Y_DIM)
 244             {
 245                 _simd_store_ps(pfBuf, valZ);
 246                 pfBuf += KNOB_SIMD_WIDTH;
 247             }
 248         }
 249     }
 250 }
 251
 252 void HotTileMgr::ClearStencilHotTile(const HOTTILE* pHotTile)
 253 {
 254     // convert from F32 to U8.
 255     uint8_t clearVal = (uint8_t)(pHotTile->clearData[0]);
 256     //broadcast 32x into __m256i...
 257     simdscalari valS = _simd_set1_epi8(clearVal);
 258
 259     simdscalari* pBuf = (simdscalari*)pHotTile->pBuffer;
 260     uint32_t numSamples = pHotTile->numSamples;
 261
 262     for (uint32_t row = 0; row < KNOB_MACROTILE_Y_DIM; row += KNOB_TILE_Y_DIM)
 263     {
 264         for (uint32_t col = 0; col < KNOB_MACROTILE_X_DIM; col += KNOB_TILE_X_DIM)
 265         {
 266             // We're putting 4 pixels in each of the 32-bit slots, so increment 4 times as quickly.
 267             for (uint32_t si = 0; si < (KNOB_TILE_X_DIM * KNOB_TILE_Y_DIM * numSamples); si += SIMD_TILE_X_DIM * SIMD_TILE_Y_DIM * 4)
 268             {
 269                 _simd_store_si(pBuf, valS);
 270                 pBuf += 1;
 271             }
 272         }
 273     }
 274 }
 275
 276 //////////////////////////////////////////////////////////////////////////
 277 /// @brief InitializeHotTiles
 278 /// for draw calls, we initialize the active hot tiles and perform deferred
 279 /// load on them if tile is in invalid state. we do this in the outer thread
 280 /// loop instead of inside the draw routine itself mainly for performance,
 281 /// to avoid unnecessary setup every triangle
 282 /// @todo support deferred clear
 283 /// @param pCreateInfo - pointer to creation info.
 284 void HotTileMgr::InitializeHotTiles(SWR_CONTEXT* pContext, DRAW_CONTEXT* pDC, uint32_t workerId, uint32_t macroID)
 285 {
 286     const API_STATE& state = GetApiState(pDC);
 287
 288     uint32_t x, y;
 289     MacroTileMgr::getTileIndices(macroID, x, y);
 290     x *= KNOB_MACROTILE_X_DIM;
 291     y *= KNOB_MACROTILE_Y_DIM;
 292
 293     uint32_t numSamples = GetNumSamples(state.rastState.sampleCount);
 294
 295     // check RT if enabled
 296     unsigned long rtSlot = 0;
 297     uint32_t colorHottileEnableMask = state.colorHottileEnable;
 298     while (_BitScanForward(&rtSlot, colorHottileEnableMask))
 299     {
 300         HOTTILE* pHotTile = GetHotTile(pContext, pDC, macroID, (SWR_RENDERTARGET_ATTACHMENT)(SWR_ATTACHMENT_COLOR0 + rtSlot), true, numSamples);
 301
 302         if (pHotTile->state == HOTTILE_INVALID)
 303         {
 304             AR_BEGIN(BELoadTiles, pDC->drawId);
 305             // invalid hottile before draw requires a load from surface before we can draw to it
 306             pContext->pfnLoadTile(GetPrivateState(pDC), KNOB_COLOR_HOT_TILE_FORMAT, (SWR_RENDERTARGET_ATTACHMENT)(SWR_ATTACHMENT_COLOR0 + rtSlot), x, y, pHotTile->renderTargetArrayIndex, pHotTile->pBuffer);
 307             pHotTile->state = HOTTILE_DIRTY;
 308             AR_END(BELoadTiles, 0);
 309         }
 310         else if (pHotTile->state == HOTTILE_CLEAR)
 311         {
 312             AR_BEGIN(BELoadTiles, pDC->drawId);
 313             // Clear the tile.
 314             ClearColorHotTile(pHotTile);
 315             pHotTile->state = HOTTILE_DIRTY;
 316             AR_END(BELoadTiles, 0);
 317         }
 318         colorHottileEnableMask &= ~(1 << rtSlot);
 319     }
 320
 321     // check depth if enabled
 322     if (state.depthHottileEnable)
 323     {
 324         HOTTILE* pHotTile = GetHotTile(pContext, pDC, macroID, SWR_ATTACHMENT_DEPTH, true, numSamples);
 325         if (pHotTile->state == HOTTILE_INVALID)
 326         {
 327             AR_BEGIN(BELoadTiles, pDC->drawId);
 328             // invalid hottile before draw requires a load from surface before we can draw to it
 329             pContext->pfnLoadTile(GetPrivateState(pDC), KNOB_DEPTH_HOT_TILE_FORMAT, SWR_ATTACHMENT_DEPTH, x, y, pHotTile->renderTargetArrayIndex, pHotTile->pBuffer);
 330             pHotTile->state = HOTTILE_DIRTY;
 331             AR_END(BELoadTiles, 0);
 332         }
 333         else if (pHotTile->state == HOTTILE_CLEAR)
 334         {
 335             AR_BEGIN(BELoadTiles, pDC->drawId);
 336             // Clear the tile.
 337             ClearDepthHotTile(pHotTile);
 338             pHotTile->state = HOTTILE_DIRTY;
 339             AR_END(BELoadTiles, 0);
 340         }
 341     }
 342
 343     // check stencil if enabled
 344     if (state.stencilHottileEnable)
 345     {
 346         HOTTILE* pHotTile = GetHotTile(pContext, pDC, macroID, SWR_ATTACHMENT_STENCIL, true, numSamples);
 347         if (pHotTile->state == HOTTILE_INVALID)
 348         {
 349             AR_BEGIN(BELoadTiles, pDC->drawId);
 350             // invalid hottile before draw requires a load from surface before we can draw to it
 351             pContext->pfnLoadTile(GetPrivateState(pDC), KNOB_STENCIL_HOT_TILE_FORMAT, SWR_ATTACHMENT_STENCIL, x, y, pHotTile->renderTargetArrayIndex, pHotTile->pBuffer);
 352             pHotTile->state = HOTTILE_DIRTY;
 353             AR_END(BELoadTiles, 0);
 354         }
 355         else if (pHotTile->state == HOTTILE_CLEAR)
 356         {
 357             AR_BEGIN(BELoadTiles, pDC->drawId);
 358             // Clear the tile.
 359             ClearStencilHotTile(pHotTile);
 360             pHotTile->state = HOTTILE_DIRTY;
 361             AR_END(BELoadTiles, 0);
 362         }
 363     }
 364 }