[Polly][GSoC 2016]New API to check vectorization legality of a loop

[cs14mtech11017]

Summary

Provide an API isVectorizable(Loop *L, int *VF) to the PolyhedralInfo interface to check for vectorization legality of a loop.
If loop is not trivially vectorizable and the minimum dependence distance is greater than the input VF, update the VF with the minimum dependence distance of the loop.
Currently, the interface checks for only constant dependence distances.
We will add support for parametric distance in the future.

Depends on D21486

[jdoerfert]

include/polly/PolyhedralInfo.h

52

factor

55

Describe the update of @p VF in more details somewhere.

lib/Analysis/PolyhedralInfo.cpp

128–129

I don;t think this is necessary.

132

if checkParallel retuirns true, the loop is trivially vectorizable for all vector widths. Only if it is not parallel you have to check if MinDistancePtr is set and if so you can check if it is constant or whatever.

[cs14mtech11017]

Updated as per review comments.

lib/Analysis/PolyhedralInfo.cpp

132

I agree. Updated accordingly.

[jdoerfert]

include/polly/PolyhedralInfo.h

52

You do not update VF to a bigger value, do you? If not remove that sentence.

55

Because you update VF in certain cases you have to describe what happens exactly for which case and that should make sense. For example you should set VF to a "special" value if the loop is trivially vectorizable because otherwise there is no difference to the case where VF is the biggest vector with possible.

lib/Analysis/PolyhedralInfo.cpp

112–113

We try to use a static_cast<int *>(User) for these cases.

118

Why do we need a cast here [what's the return type of get_num_si]? Is int really what we want or wouldn't unsigned or uint_64 be better for the minimal dependence distance?

136

Please check explicitly fi MinDistancePtr is null even if the isl functions "do the right thing" in that case.

139

999 is an odd value... maybe use INT_MAX or *VF?

[Meinersbur]

include/polly/PolyhedralInfo.h

48–49

I am a but confused with the terminology. What is the difference between "dependence distance" and "vectorization factor? Do they have the same meaning as "SIMD width"?

52

If I read correctly, it updates VF only to bigger values (and to at most 999). In case of smaller VF, the function would return false ("Not vectorizable with that VF"). Strangely, it is not updated if the VF is "infinite" (parallel).

lib/Analysis/PolyhedralInfo.cpp

71

Could we also print the max vectorization factor?

[jdoerfert]

include/polly/PolyhedralInfo.h

48–49

"dependence distance" ==> The distance (in loop iterations) of a (loop carried) dependence.
"vectorization factor" == "SIMD width"

We use the same terminology in Polly already (Ast and NodeBuilder).

52

I agree and I tried to state these problems earlier already.

lib/Analysis/PolyhedralInfo.cpp

142

There is no "min computation" happening in the getMinDependenceDistance function. It will just update MinDepDistance and the distance of the last dependence will be the resutl...

[cs14mtech11017]

include/polly/PolyhedralInfo.h

52

We do update VF to a bigger value.
If the loop is trivially vectorizable, set VF to INT_MAX. Thanks Michael for pointing that out.
If the loop is vectorizable for the given VF, update VF to the minimum dependence distance, which is the permissible VF. User needs to check for power of 2.

If the minimum dependence distance is not constant, loop is not vectorizable and do not update VF. We will work on non constant distance later at some point.

lib/Analysis/PolyhedralInfo.cpp

118

isl_val_get_num_si return a long. Not sure whether we will actually work on such a big value.

142

I am not sure what you meant by that.
It is actually computing the minimum dependence distance. Please refer to test case test/Isl/Ast/dependence_distance_varying_multiple.ll.

[jdoerfert]

In general this LGTM. Maybe look at my last comments and once you finalize this patch ask somebody to push it for you.

include/polly/PolyhedralInfo.h

49

compute the maximum vectorization factor (as that is what is described below).

52

Do we really need to provide a initial VF or couldn't we just say it is updated by the function accordingly? Then it will return false if not vectorizable, true if it is and VF will be initialized to the minimal dependence distance = maximal vectorization factor = maximal simd widht.

58

remove "give" as it might have been updated.

lib/Analysis/PolyhedralInfo.cpp

142

I think I was misstaken anyway. Just forget about it.

include/polly/PolyhedralInfo.h

   /// @return  Returns true, if loop is parallel false otherwise.
   bool isParallel(llvm::Loop *L) const;
 
+  /// @brief   Check if a given loop is vectorizable and compute the minimum
+  ///          dependence distance.
+  ///
+  /// @param L  The loop.
+  /// @param VF The vectorization factor. Update if bigger VF possible.
+  ///           If loop is trivially vectorizable, set VF to INT_MAX.
+  ///           If loop is vectorizable for the given VF, update the VF to the
+  ///           minimum dependence distance, which is the maximum permissible
+  ///           VF.
+  ///
+  /// @return  Returns true if loop is vectorizable for the given @p VF, false
+  ///          otherwise.
+  bool isVectorizable(llvm::Loop *L, int *VF) const;
+
   /// @brief   Return the SCoP containing the @p L loop.
   ///
   /// @param L The loop.

lib/Analysis/PolyhedralInfo.cpp

 
 void PolyhedralInfo::print(raw_ostream &OS, const Module *) const {
   auto &LI = getAnalysis<LoopInfoWrapperPass>().getLoopInfo();
+  int VF = 2;
   for (auto *TopLevelLoop : LI) {
     for (auto *L : depth_first(TopLevelLoop)) {
       OS.indent(2) << L->getHeader()->getName() << ":\t";

         OS << "Loop is parallel.\n";
       else if (CheckParallel)
         OS << "Loop is not parallel.\n";
+      if (CheckVectorizable && isVectorizable(L, &VF)) {
+        OS << "Loop is vectorizable with max VF = ";
+        if (VF == INT_MAX)
+          OS << "infinite";
+        else
+          OS << VF;
+        OS << "\n";
+      } else if (CheckVectorizable)
+        OS << "Loop is not vectorizable.\n";
     }
   }
 }

 
 bool PolyhedralInfo::isParallel(Loop *L) const { return checkParallel(L); }
 
+/// @brief Get the minimum dependence distance from the quasi-affine expression
+///        @p Dist and update it to @p User
+static isl_stat getMinDependenceDistance(__isl_take isl_set *Set,
+                                         __isl_take isl_aff *Dist, void *User) {
+  int *MinDepDistance = static_cast<int *>(User);
+
+  isl_val *Distance = isl_aff_get_constant_val(Dist);
+  int Flag = isl_val_cmp_si(Distance, *MinDepDistance);
+  if (Flag == -1)
+    *MinDepDistance = isl_val_get_num_si(Distance);
+
+  isl_val_free(Distance);
+  isl_set_free(Set);
+  isl_aff_free(Dist);
+  return isl_stat_ok;
+}
+
+bool PolyhedralInfo::isVectorizable(Loop *L, int *VF) const {
+  isl_pw_aff *MinDistancePtr = nullptr;
+  bool IsVectorizable = checkParallel(L, &MinDistancePtr);
+
+  if (IsVectorizable || !MinDistancePtr) {
+    // Trivially Vectorizable. Set VF to infinity
+    *VF = IsVectorizable ? INT_MAX : *VF;
+    isl_pw_aff_free(MinDistancePtr);
+    return IsVectorizable;
+  }
+
+  // Currently handle only constant distances
+  if (!isl_pw_aff_is_empty(MinDistancePtr) &&
+      !isl_pw_aff_involves_nan(MinDistancePtr) &&
+      isl_pw_aff_is_cst(MinDistancePtr)) {
+    int MinDepDistance = INT_MAX;
+
+    isl_pw_aff_foreach_piece(MinDistancePtr, getMinDependenceDistance,
+                             &MinDepDistance);
+    IsVectorizable = (*VF <= MinDepDistance) ? true : false;
+
+    // If minimum dependence distance is greater than vectorization factor @p VF
+    // provided by user, update VF to maximum permissible value, which is the
+    // minimum dependence distance.
+    *VF = IsVectorizable ? MinDepDistance : *VF;
+    DEBUG(dbgs() << "Min Dep Distance(constant):\t" << MinDepDistance << "\n");
+  }
+
+  isl_pw_aff_free(MinDistancePtr);
+  return IsVectorizable;
+}
+
 const Scop *PolyhedralInfo::getScopContainingLoop(Loop *L) const {
   assert((SI) && "ScopInfoWrapperPass is required by PolyhedralInfo pass!\n");
   for (auto &It : *SI) {

test/Isl/Ast/OpenMP/multiple_loops_outer_parallel.ll

 ; RUN: opt %loadPolly -polly-ast -polly-parallel -polly-parallel-force -analyze < %s | FileCheck %s
 ; RUN: opt %loadPolly -polyhedral-info -polly-check-parallel -analyze < %s | FileCheck %s -check-prefix=PINFO
+; RUN: opt %loadPolly -polyhedral-info -polly-check-vectorizable -analyze < %s | FileCheck %s -check-prefix=PINFO-VEC
 ;
 ;       void jd(int *A) {
 ; CHECK:  #pragma omp parallel for
 ; PINFO:  for.cond2: Loop is parallel.
+; PINFO-VEC:  for.cond2: Loop is vectorizable with max VF = infinite
 ;         for (int i = 0; i < 1024; i++)
 ;           A[i] = 1;
 ; CHECK:  #pragma omp parallel for
 ; PINFO:  for.cond: Loop is parallel.
+; PINFO-VEC:  for.cond: Loop is vectorizable with max VF = infinite
 ;         for (int i = 0; i < 1024; i++)
 ;           A[i] = A[i] * 2;
 ;       }

test/Isl/Ast/dependence_distance_constant.ll

 ; RUN: opt %loadPolly -polly-ast -polly-ast-detect-parallel -analyze < %s | FileCheck %s
 ; RUN: opt %loadPolly -polyhedral-info -polly-check-parallel -analyze < %s | FileCheck %s -check-prefix=PINFO
+; RUN: opt %loadPolly -polyhedral-info -polly-check-vectorizable -analyze < %s | FileCheck %s -check-prefix=PINFO-VEC
 ;
 ;        void f(int *A, int N) {
 ; CHECK:   #pragma minimal dependence distance: 1
 ; PINFO:   for.cond: Loop is not parallel.
+; PINFO-VEC:   for.cond: Loop is not vectorizable.
 ;          for (int j = 0; j < N; j++)
 ; CHECK:      #pragma minimal dependence distance: 8
 ; PINFO-NEXT: for.cond1: Loop is not parallel.
+; PINFO-VEC:  for.cond1: Loop is vectorizable with max VF = 8
 ;             for (int i = 0; i < N; i++)
 ;               A[i + 8] = A[i] + 1;
 ;        }

test/Isl/Ast/dependence_distance_multiple_constant.ll

 ; RUN: opt %loadPolly -basicaa -polly-ast -polly-ast-detect-parallel -analyze < %s | FileCheck %s
 ; RUN: opt %loadPolly -polyhedral-info -polly-check-parallel -analyze < %s | FileCheck %s -check-prefix=PINFO
+; RUN: opt %loadPolly -polyhedral-info -polly-check-vectorizable -analyze < %s | FileCheck %s -check-prefix=PINFO-VEC
 ;
 ;        void f(int *restrict A, int *restrict B, int N) {
 ; CHECK:   #pragma minimal dependence distance: 5
 ; PINFO:   for.cond: Loop is not parallel.
+; PINFO-VEC:   for.cond: Loop is vectorizable with max VF = 5 
 ;          for (int i = 0; i < N; i++) {
 ;            A[i + 7] = A[i] + 1;
 ;            B[i + 5] = B[i] + 1;

test/Isl/Ast/dependence_distance_varying.ll

 ; RUN: opt %loadPolly -polly-ast -polly-ast-detect-parallel -analyze < %s | FileCheck %s
 ; RUN: opt %loadPolly -polyhedral-info -polly-check-parallel -analyze < %s | FileCheck %s -check-prefix=PINFO
+; RUN: opt %loadPolly -polyhedral-info -polly-check-vectorizable -analyze < %s | FileCheck %s -check-prefix=PINFO-VEC
 ;
 ;         void f(int *A, int N) {
 ; CHECK:    #pragma minimal dependence distance: ((N - 1) % 2) + 1
 ; PINFO:    for.cond: Loop is not parallel.
+; PINFO-VEC: for.cond: Loop is not vectorizable.
 ;           for (int i = 0; i < N; i++)
 ;             A[i] = A[N - i] + 1;
 ;         }

test/Isl/Ast/dependence_distance_varying_multiple.ll

 ; RUN: opt %loadPolly -basicaa -polly-ast -polly-ast-detect-parallel -analyze < %s | FileCheck %s
 ; RUN: opt %loadPolly -polyhedral-info -polly-check-parallel -analyze < %s | FileCheck %s -check-prefix=PINFO
+; RUN: opt %loadPolly -polyhedral-info -polly-check-vectorizable -analyze < %s | FileCheck %s -check-prefix=PINFO-VEC
 ;
 ;        void f(int *restrict A, int *restrict B, int *restrict C, int *restrict D,
 ;               int *restrict E, int N) {
 ; CHECK:   #pragma minimal dependence distance: N >= 35 ? 1 : N >= 17 && N <= 34 ? 2 : 5
 ; PINFO:   for.cond: Loop is not parallel.
+; PINFO-VEC: for.cond: Loop is not vectorizable.
 ;          for (int i = 0; i < N; i++) {
 ;            A[i] = A[100 - 2 * i] + 1;
 ;            B[i] = B[100 - 3 * i] + 1;